Skip to content
New issue

Have a question about this project? Sign up for a free GitHub account to open an issue and contact its maintainers and the community.

Sign up for GitHub

By clicking “Sign up for GitHub”, you agree to our terms of service and privacy statement. We’ll occasionally send you account related emails.

Already on GitHub? Sign in to your account

CarpetX: avoid features not supported by nvhpc's compiler #331

Open
wants to merge 2 commits into
base: main
Choose a base branch
from
Open

CarpetX: avoid features not supported by nvhpc's compiler #331

Show file tree
Hide file tree
Changes from all commits
Commits
Show all changes
2 commits
Select commit Hold shift + click to select a range
45e7ade
CarpetX: manually implement custom OpenMP reductions for nvc++
rhaas80 Jan 10, 2025
017629a
TestArrayGroup: avoid use of non-standard RANK function
rhaas80 Jan 10, 2025
File filter

Filter by extension

Filter by extension
Conversations
Failed to load comments.
Loading
Jump to
Jump to file
Failed to load files.
Loading
Diff view
Diff view
70 changes: 41 additions & 29 deletions CarpetX/src/reduction.cxx
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -201,38 +201,50 @@ reduction<CCTK_REAL, dim> reduce(int gi, int vi, int tl) {
// TODO: check that multi-threading actually helps (and we are
// not dominated by memory latency)
// TODO: document required version of OpenMP to use custom reductions
#ifdef __NVCOMPILER
#pragma omp parallel
{
auto &outer = red;
reduction<CCTK_REAL, dim> red;
#else
#pragma omp parallel reduction(reduction : red)
for (amrex::MFIter mfi(mfab, mfitinfo); mfi.isValid(); ++mfi) {
const amrex::Box &bx = mfi.tilebox(); // current tile (without ghosts)
const vect<int, dim> tmin{bx.smallEnd(0), bx.smallEnd(1),
bx.smallEnd(2)};
const vect<int, dim> tmax{bx.bigEnd(0) + 1, bx.bigEnd(1) + 1,
bx.bigEnd(2) + 1};
const amrex::Box &vbx =
mfi.validbox(); // interior region (without ghosts)
const vect<int, dim> imin{vbx.smallEnd(0), vbx.smallEnd(1),
vbx.smallEnd(2)};
const vect<int, dim> imax{vbx.bigEnd(0) + 1, vbx.bigEnd(1) + 1,
vbx.bigEnd(2) + 1};

const amrex::Array4<const CCTK_REAL> &vars = mfab.array(mfi);

std::unique_ptr<amrex::Array4<const int> > finemask;
if (finemask_imfab) {
finemask = make_unique<amrex::Array4<const int> >(
finemask_imfab->array(mfi));
// Ensure the mask has the correct size
assert(finemask->begin.x == vars.begin.x);
assert(finemask->begin.y == vars.begin.y);
assert(finemask->begin.z == vars.begin.z);
assert(finemask->end.x == vars.end.x);
assert(finemask->end.y == vars.end.y);
assert(finemask->end.z == vars.end.z);
#endif
for (amrex::MFIter mfi(mfab, mfitinfo); mfi.isValid(); ++mfi) {
const amrex::Box &bx = mfi.tilebox(); // current tile (without ghosts)
const vect<int, dim> tmin{bx.smallEnd(0), bx.smallEnd(1),
bx.smallEnd(2)};
const vect<int, dim> tmax{bx.bigEnd(0) + 1, bx.bigEnd(1) + 1,
bx.bigEnd(2) + 1};
const amrex::Box &vbx =
mfi.validbox(); // interior region (without ghosts)
const vect<int, dim> imin{vbx.smallEnd(0), vbx.smallEnd(1),
vbx.smallEnd(2)};
const vect<int, dim> imax{vbx.bigEnd(0) + 1, vbx.bigEnd(1) + 1,
vbx.bigEnd(2) + 1};

const amrex::Array4<const CCTK_REAL> &vars = mfab.array(mfi);

std::unique_ptr<amrex::Array4<const int> > finemask;
if (finemask_imfab) {
finemask = make_unique<amrex::Array4<const int> >(
finemask_imfab->array(mfi));
// Ensure the mask has the correct size
assert(finemask->begin.x == vars.begin.x);
assert(finemask->begin.y == vars.begin.y);
assert(finemask->begin.z == vars.begin.z);
assert(finemask->end.x == vars.end.x);
assert(finemask->end.y == vars.end.y);
assert(finemask->end.z == vars.end.z);
}

red += reduce_array(vars, vi, tmin, tmax, indextype, imin, imax,
finemask.get(), x0, dx);
}

red += reduce_array(vars, vi, tmin, tmax, indextype, imin, imax,
finemask.get(), x0, dx);
#ifdef __NVCOMPILER
#pragma omp critical
outer += red;
}
#endif
}
}

Expand Down
2 changes: 2 additions & 0 deletions CarpetX/src/reduction.hxx
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -121,7 +121,9 @@ template <typename T, int D> struct combine {
};

typedef reduction<CCTK_REAL, dim> reduction_CCTK_REAL;
#ifndef __NVCOMPILER
#pragma omp declare reduction(reduction:reduction_CCTK_REAL : omp_out += omp_in)
#endif

MPI_Datatype reduction_mpi_datatype_CCTK_REAL();
MPI_Op reduction_mpi_op();
Expand Down
22 changes: 9 additions & 13 deletions TestArrayGroup/src/TestDynamicData.F90
View file
Open in desktop
Original file line number Diff line number Diff line change
Expand Up @@ -6,25 +6,21 @@ subroutine TestArrayGroup_DynamicDataF(CCTK_ARGUMENTS)
DECLARE_CCTK_PARAMETERS
DECLARE_CCTK_ARGUMENTS

! Validate grid array dynamic data
if(RANK(test1) /= 3) then ! note rank 3 b/c of vector of rank=2 arrays
call CCTK_ERROR("incorrect dimension in test1 array dynamic data")
endif
if(SIZE(test1, 1) /= 5 .or. SIZE(test1, 2) /= 6 .or. SIZE(test1, 3) /= 4) then
integer, dimension(3) :: sizes1, sizes2, sizes3

! check that grid variable is of rank 3. This fails to compiler otherwise.
sizes1 = SHAPE(test1)
if(sizes1(1) /= 5 .or. sizes1(2) /= 6 .or. sizes1(3) /= 4) then
call CCTK_ERROR("incorrect size in test1 array dynamic data")
endif

if(RANK(test2) /= 3) then ! note rank 3 b/c of vector of rank=2 arrays
call CCTK_ERROR("incorrect dimension in test2 array dynamic data")
endif
if(SIZE(test2, 1) /= 5 .or. SIZE(test2, 2) /= 6 .or. SIZE(test2, 3) /= 4) then
sizes2 = SHAPE(test2)
if(sizes2(1) /= 5 .or. sizes2(2) /= 6 .or. sizes2(3) /= 4) then
call CCTK_ERROR("incorrect size in test2 array dynamic data")
endif

if(RANK(test3) /= 3) then ! note rank 3 b/c of vector of rank=2 arrays
call CCTK_ERROR("incorrect dimension in test3 array dynamic data")
endif
if(SIZE(test3, 1) /= 5 .or. SIZE(test3, 2) /= 6 .or. SIZE(test3, 3) /= 4) then
sizes3 = SHAPE(test3)
if(sizes3(1) /= 5 .or. sizes3(2) /= 6 .or. sizes3(3) /= 4) then
call CCTK_ERROR("incorrect size in test3 array dynamic data")
endif

Expand Down
Loading