Skip to content

Commit f529add

Browse files
jeremyltjeremylt
authored
Merge pull request #1688 from CEED/jeremy/cuda-asan
GPU CI Updates
2 parents dfc3c7d + f329610 commit f529add

File tree

7 files changed

+59
-31
lines changed

7 files changed

+59
-31
lines changed

‎.gitlab-ci.yml‎

Lines changed: 23 additions & 15 deletions
Original file line numberDiff line numberDiff line change
@@ -186,6 +186,10 @@ noether-cuda:
186186
- echo "-------------- FC ------------------" && $FC --version
187187
- echo "-------------- NVCC ----------------" && $NVCC --version
188188
- echo "-------------- GCOV ----------------" && gcov --version
189+
# ASAN
190+
- echo "-------------- ASAN ----------------"
191+
- export ASAN=1 AFLAGS="-fsanitize=address -fsanitize=leak" ASAN_OPTIONS=protect_shadow_gap=0
192+
- echo $AFLAGS
189193
script:
190194
- rm -f .SUCCESS
191195
# libCEED
@@ -200,11 +204,29 @@ noether-cuda:
200204
# Note: PETSC_DIR is set by default in GitLab runner env, unsetting to isolate core tests
201205
- export PETSC_DIR= PETSC_ARCH=
202206
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="cuda" junit realsearch=%
207+
# Rebuild without ASAN
208+
- unset ASAN AFLAGS ASAN_OPTIONS
209+
- make clean
210+
- PEDANTIC=1 make -k -j$NPROC_CPU -l$NPROC_CPU
203211
# Libraries for examples
204212
# -- PETSc with CUDA (minimal)
205213
- export PETSC_DIR=/projects/petsc PETSC_ARCH=mpich-cuda-O PETSC_OPTIONS='-use_gpu_aware_mpi 0' && git -C $PETSC_DIR -c safe.directory=$PETSC_DIR describe
206214
- echo "-------------- PETSc ---------------" && make -C $PETSC_DIR info
207-
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="cuda" junit search="petsc fluids-navierstokes solids"
215+
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="cuda" junit search="petsc fluids solids"
216+
# -- MFEM v4.7
217+
- cd .. && export MFEM_VERSION=mfem-4.7 && { [[ -d $MFEM_VERSION ]] || { git clone --depth 1 --branch v4.7 https://github.com/mfem/mfem.git $MFEM_VERSION && make -C $MFEM_VERSION -j$(nproc) serial CXXFLAGS="-O -std=c++11"; }; } && export MFEM_DIR=$PWD/$MFEM_VERSION && cd libCEED
218+
- echo "-------------- MFEM ----------------" && make -C $MFEM_DIR info
219+
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="cuda" junit search=mfem
220+
# -- Nek5000 v19.0
221+
- export COVERAGE=0
222+
- cd .. && export NEK5K_VERSION=Nek5000-19.0 && { [[ -d $NEK5K_VERSION ]] || { git clone --depth 1 --branch v19.0 https://github.com/Nek5000/Nek5000.git $NEK5K_VERSION && cd $NEK5K_VERSION/tools && ./maketools genbox genmap reatore2 && cd ../..; }; } && export NEK5K_DIR=$PWD/$NEK5K_VERSION && export PATH=$NEK5K_DIR/bin:$PATH MPI=0 && cd libCEED
223+
- echo "-------------- Nek5000 -------------" && git -C $NEK5K_DIR describe --tags
224+
- export NPROC_POOL=1
225+
- make -k -j$NPROC_GPU BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="cuda" junit search=nek NEK5K_DIR=$NEK5K_DIR
226+
# -- deal.II 8bd5c262f13e15793aa206b6eed8774a9b25ce11
227+
- export DEAL_II_ROOT_DIR=/projects/dealii DEAL_II_DIR=/projects/dealii/install
228+
- echo "-------------- deal.II -------------" && git -C $DEAL_II_ROOT_DIR -c safe.directory=$DEAL_II_ROOT_DIR describe --always
229+
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="cuda" junit search=dealii DEAL_II_DIR=$DEAL_II_DIR
208230
# Clang-tidy
209231
- echo "-------------- clang-tidy ----------" && clang-tidy --version
210232
- TIDY_OPTS="-fix-errors" make -j$NPROC_CPU tidy && git diff --color=always --exit-code
@@ -269,20 +291,6 @@ noether-rocm:
269291
- export PETSC_DIR=/projects/petsc PETSC_ARCH=mpich-hip && git -C $PETSC_DIR -c safe.directory=$PETSC_DIR describe
270292
- echo "-------------- PETSc ---------------" && make -C $PETSC_DIR info
271293
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="hip" junit search="petsc fluids solids"
272-
# -- MFEM v4.7
273-
- cd .. && export MFEM_VERSION=mfem-4.7 && { [[ -d $MFEM_VERSION ]] || { git clone --depth 1 --branch v4.7 https://github.com/mfem/mfem.git $MFEM_VERSION && make -C $MFEM_VERSION -j$(nproc) serial CXXFLAGS="-O -std=c++11"; }; } && export MFEM_DIR=$PWD/$MFEM_VERSION && cd libCEED
274-
- echo "-------------- MFEM ----------------" && make -C $MFEM_DIR info
275-
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="hip" junit search=mfem
276-
# -- Nek5000 v19.0
277-
- export COVERAGE=0
278-
- cd .. && export NEK5K_VERSION=Nek5000-19.0 && { [[ -d $NEK5K_VERSION ]] || { git clone --depth 1 --branch v19.0 https://github.com/Nek5000/Nek5000.git $NEK5K_VERSION && cd $NEK5K_VERSION/tools && ./maketools genbox genmap reatore2 && cd ../..; }; } && export NEK5K_DIR=$PWD/$NEK5K_VERSION && export PATH=$NEK5K_DIR/bin:$PATH MPI=0 && cd libCEED
279-
- echo "-------------- Nek5000 -------------" && git -C $NEK5K_DIR describe --tags
280-
- export NPROC_POOL=1
281-
- make -k -j$NPROC_GPU BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="hip" junit search=nek NEK5K_DIR=$NEK5K_DIR
282-
# -- deal.II 8bd5c262f13e15793aa206b6eed8774a9b25ce11
283-
- export DEAL_II_ROOT_DIR=/projects/dealii DEAL_II_DIR=/projects/dealii/install
284-
- echo "-------------- deal.II -------------" && git -C $DEAL_II_ROOT_DIR -c safe.directory=$DEAL_II_ROOT_DIR describe --always
285-
- make -k -j$((NPROC_GPU / NPROC_POOL)) BACKENDS="$BACKENDS_GPU" JUNIT_BATCH="hip" junit search=dealii DEAL_II_DIR=$DEAL_II_DIR
286294
# Clang-tidy
287295
- echo "-------------- clang-tidy ----------" && clang-tidy --version
288296
- TIDY_OPTS="-fix-errors" make -j$NPROC_CPU tidy && git diff --color=always --exit-code

‎backends/cuda-ref/ceed-cuda-ref-operator.c‎

Lines changed: 13 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -600,6 +600,7 @@ static int CeedOperatorSetupAtPoints_Cuda(CeedOperator op) {
600600
CeedCallBackend(CeedElemRestrictionGetNumPointsInElement(rstr_points, e, &num_points_elem));
601601
impl->num_points[e] = num_points_elem;
602602
}
603+
CeedCallBackend(CeedElemRestrictionDestroy(&rstr_points));
603604
}
604605
impl->max_num_points = max_num_points;
605606

@@ -779,6 +780,8 @@ static int CeedOperatorApplyAddAtPoints_Cuda(CeedOperator op, CeedVector in_vec,
779780
CeedCallBackend(CeedOperatorAtPointsGetPoints(op, &rstr_points, &point_coords));
780781
CeedCallBackend(CeedElemRestrictionCreateVector(rstr_points, NULL, &impl->point_coords_elem));
781782
CeedCallBackend(CeedElemRestrictionApply(rstr_points, CEED_NOTRANSPOSE, point_coords, impl->point_coords_elem, request));
783+
CeedCallBackend(CeedVectorDestroy(&point_coords));
784+
CeedCallBackend(CeedElemRestrictionDestroy(&rstr_points));
782785
}
783786

784787
// Process inputs
@@ -1538,11 +1541,9 @@ static int CeedSingleOperatorAssembleSetup_Cuda(CeedOperator op, CeedInt use_cee
15381541
CeedCallCuda(ceed, cudaMemcpy(&asmb->d_B_in[i * elem_size_in * num_qpts_in], h_B_in, elem_size_in * num_qpts_in * sizeof(CeedScalar),
15391542
cudaMemcpyHostToDevice));
15401543
}
1541-
1542-
if (identity) {
1543-
CeedCallBackend(CeedFree(&identity));
1544-
}
1544+
CeedCallBackend(CeedFree(&identity));
15451545
}
1546+
CeedCallBackend(CeedFree(&eval_modes_in));
15461547

15471548
// Load into B_out, in order that they will be used in eval_modes_out
15481549
{
@@ -1575,11 +1576,9 @@ static int CeedSingleOperatorAssembleSetup_Cuda(CeedOperator op, CeedInt use_cee
15751576
CeedCallCuda(ceed, cudaMemcpy(&asmb->d_B_out[i * elem_size_out * num_qpts_out], h_B_out, elem_size_out * num_qpts_out * sizeof(CeedScalar),
15761577
cudaMemcpyHostToDevice));
15771578
}
1578-
1579-
if (identity) {
1580-
CeedCallBackend(CeedFree(&identity));
1581-
}
1579+
CeedCallBackend(CeedFree(&identity));
15821580
}
1581+
CeedCallBackend(CeedFree(&eval_modes_out));
15831582
return CEED_ERROR_SUCCESS;
15841583
}
15851584

@@ -1743,6 +1742,8 @@ static int CeedOperatorLinearAssembleAddDiagonalAtPoints_Cuda(CeedOperator op, C
17431742
CeedCallBackend(CeedOperatorAtPointsGetPoints(op, &rstr_points, &point_coords));
17441743
CeedCallBackend(CeedElemRestrictionCreateVector(rstr_points, NULL, &impl->point_coords_elem));
17451744
CeedCallBackend(CeedElemRestrictionApply(rstr_points, CEED_NOTRANSPOSE, point_coords, impl->point_coords_elem, request));
1745+
CeedCallBackend(CeedVectorDestroy(&point_coords));
1746+
CeedCallBackend(CeedElemRestrictionDestroy(&rstr_points));
17461747
}
17471748

17481749
// Process inputs
@@ -1933,6 +1934,10 @@ static int CeedOperatorLinearAssembleAddDiagonalAtPoints_Cuda(CeedOperator op, C
19331934
for (CeedInt i = 0; i < num_input_fields; i++) {
19341935
CeedCallBackend(CeedOperatorInputRestore_Cuda(op_input_fields[i], qf_input_fields[i], i, NULL, NULL, true, impl));
19351936
}
1937+
1938+
// Restore work vector
1939+
CeedCallBackend(CeedRestoreWorkVector(ceed, &active_e_vec_in));
1940+
CeedCallBackend(CeedRestoreWorkVector(ceed, &active_e_vec_out));
19361941
return CEED_ERROR_SUCCESS;
19371942
}
19381943

‎backends/cuda-ref/ceed-cuda-ref-qfunction.c‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -68,6 +68,7 @@ static int CeedQFunctionDestroy_Cuda(CeedQFunction qf) {
6868
CeedQFunction_Cuda *data;
6969

7070
CeedCallBackend(CeedQFunctionGetData(qf, &data));
71+
CeedCallBackend(CeedFree(&data->qfunction_source));
7172
if (data->module) CeedCallCuda(CeedQFunctionReturnCeed(qf), cuModuleUnload(data->module));
7273
CeedCallBackend(CeedFree(&data));
7374
return CEED_ERROR_SUCCESS;

‎backends/cuda-ref/ceed-cuda-ref-restriction.c‎

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -80,6 +80,10 @@ static inline int CeedElemRestrictionSetupCompile_Cuda(CeedElemRestriction rstr)
8080
"USE_DETERMINISTIC", is_deterministic ? 1 : 0));
8181
CeedCallBackend(CeedGetKernel_Cuda(ceed, impl->module, "OffsetNoTranspose", &impl->ApplyNoTranspose));
8282
CeedCallBackend(CeedGetKernel_Cuda(ceed, impl->module, "AtPointsTranspose", &impl->ApplyTranspose));
83+
// Cleanup
84+
CeedCallBackend(CeedFree(&offset_kernel_path));
85+
for (CeedInt i = 0; i < num_file_paths; i++) CeedCallBackend(CeedFree(&file_paths[i]));
86+
CeedCallBackend(CeedFree(&file_paths));
8387
} break;
8488
case CEED_RESTRICTION_STANDARD: {
8589
CeedCallBackend(CeedGetJitAbsolutePath(ceed, "ceed/jit-source/cuda/cuda-ref-restriction-offset.h", &restriction_kernel_path));

‎backends/hip-ref/ceed-hip-ref-operator.c‎

Lines changed: 13 additions & 8 deletions
Original file line numberDiff line numberDiff line change
@@ -598,6 +598,7 @@ static int CeedOperatorSetupAtPoints_Hip(CeedOperator op) {
598598
CeedCallBackend(CeedElemRestrictionGetNumPointsInElement(rstr_points, e, &num_points_elem));
599599
impl->num_points[e] = num_points_elem;
600600
}
601+
CeedCallBackend(CeedElemRestrictionDestroy(&rstr_points));
601602
}
602603
impl->max_num_points = max_num_points;
603604

@@ -777,6 +778,8 @@ static int CeedOperatorApplyAddAtPoints_Hip(CeedOperator op, CeedVector in_vec,
777778
CeedCallBackend(CeedOperatorAtPointsGetPoints(op, &rstr_points, &point_coords));
778779
CeedCallBackend(CeedElemRestrictionCreateVector(rstr_points, NULL, &impl->point_coords_elem));
779780
CeedCallBackend(CeedElemRestrictionApply(rstr_points, CEED_NOTRANSPOSE, point_coords, impl->point_coords_elem, request));
781+
CeedCallBackend(CeedVectorDestroy(&point_coords));
782+
CeedCallBackend(CeedElemRestrictionDestroy(&rstr_points));
780783
}
781784

782785
// Process inputs
@@ -1535,11 +1538,9 @@ static int CeedSingleOperatorAssembleSetup_Hip(CeedOperator op, CeedInt use_ceed
15351538
CeedCallHip(ceed, hipMemcpy(&asmb->d_B_in[i * elem_size_in * num_qpts_in], h_B_in, elem_size_in * num_qpts_in * sizeof(CeedScalar),
15361539
hipMemcpyHostToDevice));
15371540
}
1538-
1539-
if (identity) {
1540-
CeedCallBackend(CeedFree(&identity));
1541-
}
1541+
CeedCallBackend(CeedFree(&identity));
15421542
}
1543+
CeedCallBackend(CeedFree(&eval_modes_in));
15431544

15441545
// Load into B_out, in order that they will be used in eval_modes_out
15451546
{
@@ -1572,11 +1573,9 @@ static int CeedSingleOperatorAssembleSetup_Hip(CeedOperator op, CeedInt use_ceed
15721573
CeedCallHip(ceed, hipMemcpy(&asmb->d_B_out[i * elem_size_out * num_qpts_out], h_B_out, elem_size_out * num_qpts_out * sizeof(CeedScalar),
15731574
hipMemcpyHostToDevice));
15741575
}
1575-
1576-
if (identity) {
1577-
CeedCallBackend(CeedFree(&identity));
1578-
}
1576+
CeedCallBackend(CeedFree(&identity));
15791577
}
1578+
CeedCallBackend(CeedFree(&eval_modes_out));
15801579
return CEED_ERROR_SUCCESS;
15811580
}
15821581

@@ -1740,6 +1739,8 @@ static int CeedOperatorLinearAssembleAddDiagonalAtPoints_Hip(CeedOperator op, Ce
17401739
CeedCallBackend(CeedOperatorAtPointsGetPoints(op, &rstr_points, &point_coords));
17411740
CeedCallBackend(CeedElemRestrictionCreateVector(rstr_points, NULL, &impl->point_coords_elem));
17421741
CeedCallBackend(CeedElemRestrictionApply(rstr_points, CEED_NOTRANSPOSE, point_coords, impl->point_coords_elem, request));
1742+
CeedCallBackend(CeedVectorDestroy(&point_coords));
1743+
CeedCallBackend(CeedElemRestrictionDestroy(&rstr_points));
17431744
}
17441745

17451746
// Process inputs
@@ -1930,6 +1931,10 @@ static int CeedOperatorLinearAssembleAddDiagonalAtPoints_Hip(CeedOperator op, Ce
19301931
for (CeedInt i = 0; i < num_input_fields; i++) {
19311932
CeedCallBackend(CeedOperatorInputRestore_Hip(op_input_fields[i], qf_input_fields[i], i, NULL, NULL, true, impl));
19321933
}
1934+
1935+
// Restore work vector
1936+
CeedCallBackend(CeedRestoreWorkVector(ceed, &active_e_vec_in));
1937+
CeedCallBackend(CeedRestoreWorkVector(ceed, &active_e_vec_out));
19331938
return CEED_ERROR_SUCCESS;
19341939
}
19351940

‎backends/hip-ref/ceed-hip-ref-qfunction.c‎

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -70,6 +70,7 @@ static int CeedQFunctionDestroy_Hip(CeedQFunction qf) {
7070
CeedQFunction_Hip *data;
7171

7272
CeedCallBackend(CeedQFunctionGetData(qf, &data));
73+
CeedCallBackend(CeedFree(&data->qfunction_source));
7374
if (data->module) CeedCallHip(CeedQFunctionReturnCeed(qf), hipModuleUnload(data->module));
7475
CeedCallBackend(CeedFree(&data));
7576
return CEED_ERROR_SUCCESS;

‎backends/hip-ref/ceed-hip-ref-restriction.c‎

Lines changed: 4 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -79,6 +79,10 @@ static inline int CeedElemRestrictionSetupCompile_Hip(CeedElemRestriction rstr)
7979
"USE_DETERMINISTIC", is_deterministic ? 1 : 0));
8080
CeedCallBackend(CeedGetKernel_Hip(ceed, impl->module, "OffsetNoTranspose", &impl->ApplyNoTranspose));
8181
CeedCallBackend(CeedGetKernel_Hip(ceed, impl->module, "AtPointsTranspose", &impl->ApplyTranspose));
82+
// Cleanup
83+
CeedCallBackend(CeedFree(&offset_kernel_path));
84+
for (CeedInt i = 0; i < num_file_paths; i++) CeedCallBackend(CeedFree(&file_paths[i]));
85+
CeedCallBackend(CeedFree(&file_paths));
8286
} break;
8387
case CEED_RESTRICTION_STANDARD: {
8488
CeedCallBackend(CeedGetJitAbsolutePath(ceed, "ceed/jit-source/hip/hip-ref-restriction-offset.h", &restriction_kernel_path));

0 commit comments

Comments
 (0)