diff --git a/parsec b/parsec index b3e7e24c..d0fc4c01 160000 --- a/parsec +++ b/parsec @@ -1 +1 @@ -Subproject commit b3e7e24c4ab42076ee39a520f1540a9fe6b553db +Subproject commit d0fc4c01e6d10c8cc10719c1930ce1a5677a075c diff --git a/src/cuda/lapack_cuda_stage_in.c b/src/cuda/lapack_cuda_stage_in.c index f2f48ced..5cb0a210 100644 --- a/src/cuda/lapack_cuda_stage_in.c +++ b/src/cuda/lapack_cuda_stage_in.c @@ -33,14 +33,14 @@ dplasma_cuda_lapack_stage_in(parsec_gpu_task_t *gtask, if(flow_mask & (1U << i)){ copy_in = task->data[i].data_in; copy_out = task->data[i].data_out; - ddc = (dplasma_data_collection_t*)gtask->flow_dc[i]; + ddc = (dplasma_data_collection_t*)gtask->flow_info[i].flow_dc; assert(ddc != NULL); elem_sz = parsec_datadist_getsizeoftype(ddc->dc_original->mtype); in_elem_dev = (parsec_device_gpu_module_t*)parsec_mca_device_get( copy_in->device_index); if( (in_elem_dev->super.type == PARSEC_DEV_CUDA) || (ddc->dc_original->storage != PARSEC_MATRIX_LAPACK)){ ret = (cudaError_t)cudaMemcpyAsync( copy_out->device_private, copy_in->device_private, - gtask->flow_nb_elts[i], + gtask->flow_info[i].flow_span, (in_elem_dev->super.type != PARSEC_DEV_CUDA)? cudaMemcpyHostToDevice : cudaMemcpyDeviceToDevice, cuda_stream->cuda_stream); @@ -109,7 +109,7 @@ dplasma_cuda_lapack_stage_out(parsec_gpu_task_t *gtask, if(flow_mask & (1U << i)){ copy_in = task->data[i].data_out; copy_out = copy_in->original->device_copies[0]; - ddc = (dplasma_data_collection_t*)gtask->flow_dc[i]; + ddc = (dplasma_data_collection_t*)gtask->flow_info[i].flow_dc; assert(ddc != NULL); elem_sz = parsec_datadist_getsizeoftype(ddc->dc_original->mtype); out_elem_dev = (parsec_device_gpu_module_t*)parsec_mca_device_get( copy_out->device_index); @@ -117,7 +117,7 @@ dplasma_cuda_lapack_stage_out(parsec_gpu_task_t *gtask, if( (out_elem_dev->super.type == PARSEC_DEV_CUDA) || (ddc->dc_original->storage != PARSEC_MATRIX_LAPACK)){ ret = (cudaError_t)cudaMemcpyAsync( copy_out->device_private, copy_in->device_private, - gtask->flow_nb_elts[i], + gtask->flow_info[i].flow_span, out_elem_dev->super.type != PARSEC_DEV_CUDA ? cudaMemcpyDeviceToHost : cudaMemcpyDeviceToDevice, cuda_stream->cuda_stream); diff --git a/src/dplasmaaux.c b/src/dplasmaaux.c index 5a39a41b..126456d0 100644 --- a/src/dplasmaaux.c +++ b/src/dplasmaaux.c @@ -98,7 +98,7 @@ dplasma_aux_getGEMMLookahead( parsec_tiled_matrix_t *A ) * look ahead based on the global information to get the same one on all * nodes. */ - int nbunits = vpmap_get_nb_total_threads() * A->super.nodes; + int nbunits = parsec_vpmap_get_nb_total_threads() * A->super.nodes; double alpha = 3. * (double)nbunits / ( A->mt * A->nt ); if ( A->super.nodes == 1 ) { diff --git a/src/utils/dplasma_lapack_adtt.c b/src/utils/dplasma_lapack_adtt.c index 2d47988a..14d7d7eb 100644 --- a/src/utils/dplasma_lapack_adtt.c +++ b/src/utils/dplasma_lapack_adtt.c @@ -288,7 +288,7 @@ static parsec_data_t* data_of(parsec_data_collection_t *desc, ...) "data_of CP %p [old type %p] loc %d -> dtt %p target_shape %d layout %d", cp, cp->dtt, loc, adt->opaque_dtt, info.shape, info.layout); dt = parsec_data_create_with_type( dt->dc, - dt->key, cp->device_private, dt->nb_elts, + dt->key, cp->device_private, dt->span, adt->opaque_dtt); } } diff --git a/src/zgebrd_ge2gb_wrapper.c b/src/zgebrd_ge2gb_wrapper.c index f1b93f10..404fcc3b 100644 --- a/src/zgebrd_ge2gb_wrapper.c +++ b/src/zgebrd_ge2gb_wrapper.c @@ -310,7 +310,7 @@ dplasma_zgebrd_ge2gb_New( int ib, dplasma_qrtree_t *qrtre0, *qrtree, *lqtree; int P, Q, cores; - cores = dplasma_imax( vpmap_get_nb_total_threads(), 1 ); + cores = dplasma_imax( parsec_vpmap_get_nb_total_threads(), 1 ); qrtree = malloc( sizeof(dplasma_qrtree_t) ); lqtree = malloc( sizeof(dplasma_qrtree_t) ); diff --git a/src/zgetrf_1d.jdf b/src/zgetrf_1d.jdf index 9541b481..2af38062 100644 --- a/src/zgetrf_1d.jdf +++ b/src/zgetrf_1d.jdf @@ -57,7 +57,7 @@ descIPIV [type = "parsec_tiled_matrix_t*" hidden = on default = "((dplasma_da INFO [type = "int*"] -nbmaxthrd [type = "int" hidden=on default="( dplasma_imax( 1, dplasma_imin( vpmap_get_nb_threads_in_vp(0) - 1, 48 ) ) )" ] /* 48 is the actual limit of the kernel */ +nbmaxthrd [type = "int" hidden=on default="( dplasma_imax( 1, dplasma_imin( parsec_vpmap_get_vp_threads(0) - 1, 48 ) ) )" ] /* 48 is the actual limit of the kernel */ KT [type = "int" hidden=on default="( dplasma_imin( descA->mt, descA->nt )-1 )" ] getrfdata [type = "CORE_zgetrf_data_t *" hidden = on default = "NULL"] diff --git a/src/zgetrf_1d_wrapper.c b/src/zgetrf_1d_wrapper.c index 8ce95db9..0956d26c 100644 --- a/src/zgetrf_1d_wrapper.c +++ b/src/zgetrf_1d_wrapper.c @@ -83,7 +83,7 @@ dplasma_zgetrf_1d_New( parsec_tiled_matrix_t *A, int *INFO ) { parsec_zgetrf_1d_taskpool_t *parsec_getrf_1d; - int nbthreads = dplasma_imax( 1, vpmap_get_nb_threads_in_vp(0) - 1 ); + int nbthreads = dplasma_imax( 1, parsec_vpmap_get_vp_threads(0) - 1 ); dplasma_data_collection_t * ddc_A = dplasma_wrap_data_collection((parsec_tiled_matrix_t*)A); dplasma_data_collection_t * ddc_IPIV = dplasma_wrap_data_collection((parsec_tiled_matrix_t*)IPIV); diff --git a/src/zgetrf_qrf.jdf b/src/zgetrf_qrf.jdf index 1e7ba47c..353fb362 100644 --- a/src/zgetrf_qrf.jdf +++ b/src/zgetrf_qrf.jdf @@ -52,7 +52,7 @@ INFO [type = "int*"] param_p [type = int default="((parsec_matrix_block_cyclic_t*)descA)->grid.rows" hidden=on ] param_q [type = int default="((parsec_matrix_block_cyclic_t*)descA)->grid.cols" hidden=on ] minMNT [type = int default="dplasma_imin( descA->mt-1, descA->nt-1 )" hidden=on ] -nbmaxthrd [type = "int" default="( dplasma_imax( 1, dplasma_imin( vpmap_get_nb_threads_in_vp(0) - 1, 48 ) ) )" hidden=on] /* 48 is the actual limit of the kernel */ +nbmaxthrd [type = "int" default="( dplasma_imax( 1, dplasma_imin( parsec_vpmap_get_vp_threads(0) - 1, 48 ) ) )" hidden=on] /* 48 is the actual limit of the kernel */ getrfdata [type = "CORE_zgetrf_data_t *" hidden = on default = "NULL"] diff --git a/src/zgetrf_qrf_wrapper.c b/src/zgetrf_qrf_wrapper.c index 37dc0c21..8a4c721d 100644 --- a/src/zgetrf_qrf_wrapper.c +++ b/src/zgetrf_qrf_wrapper.c @@ -166,7 +166,7 @@ dplasma_zgetrf_qrf_New( dplasma_qrtree_t *qrtree, int ib = TS->mb; size_t sizeW = 1; size_t sizeReduceVec = 1; - int nbthreads = dplasma_imax( 1, vpmap_get_nb_threads_in_vp(0) - 1 ); + int nbthreads = dplasma_imax( 1, parsec_vpmap_get_vp_threads(0) - 1 ); /* * Compute W size according to criteria used. diff --git a/tests/common.h b/tests/common.h index 5fc4befb..e5381496 100644 --- a/tests/common.h +++ b/tests/common.h @@ -22,6 +22,7 @@ #include "parsec/profiling.h" #include "parsec/parsec_internal.h" #include "parsec/utils/debug.h" +#include "parsec/interfaces/dtd/insert_function.h" #include "dplasma.h" #include "dplasma/types.h" @@ -31,6 +32,32 @@ #include "common_timing.h" #include "flops.h" +/* PaRSEC 4.1 removed parsec_dtd_create/destroy_arena_datatype helpers. */ +static inline parsec_arena_datatype_t * +dplasma_dtd_create_arena_datatype_compat(parsec_context_t *ctx, int *id) +{ + parsec_arena_datatype_t *adt = parsec_arena_datatype_new(); + if( NULL == adt ) { + return NULL; + } + if( PARSEC_SUCCESS != parsec_dtd_attach_arena_datatype(ctx, adt, id) ) { + parsec_arena_datatype_release(&adt); + return NULL; + } + return adt; +} + +static inline int +dplasma_dtd_destroy_arena_datatype_compat(parsec_context_t *ctx, int id) +{ + return parsec_dtd_free_arena_datatype(ctx, id); +} + +#define parsec_dtd_create_arena_datatype(ctx, id) \ + dplasma_dtd_create_arena_datatype_compat((ctx), (id)) +#define parsec_dtd_destroy_arena_datatype(ctx, id) \ + dplasma_dtd_destroy_arena_datatype_compat((ctx), (id)) + /* these are globals in common.c */ extern char *PARSEC_SCHED_NAME[]; extern int unix_timestamp;