From 96c86b69bf8f230c4b53503859b5db4cce24e21f Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Fri, 9 Sep 2022 13:23:00 -0700 Subject: [PATCH 1/8] Enable OpenMP in particle push and coordinate transformation routines. --- src/particles/Push.cpp | 3 +++ src/particles/transformation/CoordinateTransformation.cpp | 3 +++ 2 files changed, 6 insertions(+) diff --git a/src/particles/Push.cpp b/src/particles/Push.cpp index 828b8e3cb..10c17d571 100644 --- a/src/particles/Push.cpp +++ b/src/particles/Push.cpp @@ -118,6 +118,9 @@ namespace detail // loop over all particle boxes using ParIt = ImpactXParticleContainer::iterator; +#ifdef AMREX_USE_OMP +#pragma omp parallel if (amrex::Gpu::notInLaunchRegion()) +#endif for (ParIt pti(pc, lev); pti.isValid(); ++pti) { const int np = pti.numParticles(); //const auto t_lev = pti.GetLevel(); diff --git a/src/particles/transformation/CoordinateTransformation.cpp b/src/particles/transformation/CoordinateTransformation.cpp index e67296c00..71d4741c3 100644 --- a/src/particles/transformation/CoordinateTransformation.cpp +++ b/src/particles/transformation/CoordinateTransformation.cpp @@ -37,6 +37,9 @@ namespace transformation { for (int lev = 0; lev <= nLevel; ++lev) { // loop over all particle boxes using ParIt = ImpactXParticleContainer::iterator; +#ifdef AMREX_USE_OMP +#pragma omp parallel if (amrex::Gpu::notInLaunchRegion()) +#endif for (ParIt pti(pc, lev); pti.isValid(); ++pti) { const int np = pti.numParticles(); From ecccc14ed9b972df017c7b18f0dd85c581c26dbf Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Tue, 20 Sep 2022 11:03:09 -0700 Subject: [PATCH 2/8] enable tiling by default if not running on GPU --- src/initialization/InitParser.cpp | 13 +++++++++++++ 1 file changed, 13 insertions(+) diff --git a/src/initialization/InitParser.cpp b/src/initialization/InitParser.cpp index 604e7ef0f..7f4ee7d67 100644 --- a/src/initialization/InitParser.cpp +++ b/src/initialization/InitParser.cpp @@ -22,5 +22,18 @@ namespace impactx::initialization bool abort_on_out_of_gpu_memory = true; // AMReX' default: false pp_amrex.query("abort_on_out_of_gpu_memory", abort_on_out_of_gpu_memory); pp_amrex.add("abort_on_out_of_gpu_memory", abort_on_out_of_gpu_memory); + + // Here we override the default tiling option for particles, which is always + // "false" in AMReX, to "false" if compiling for GPU execution and "true" + // if compiling for CPU. + { + amrex::ParmParse pp_particles("particles"); +#ifdef AMREX_USE_GPU + bool do_tiling = false; // By default, tiling is off on GPU +#else + bool do_tiling = true; +#endif + pp_particles.queryAdd("do_tiling", do_tiling); + } } } // namespace impactx From 479c1e3b6549147c8c44fcd666d24ba5ee6a4c33 Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Tue, 20 Sep 2022 16:39:24 -0700 Subject: [PATCH 3/8] use dynamic scheduling by default --- src/particles/ImpactXParticleContainer.H | 26 ++++++++++++++-- src/particles/ImpactXParticleContainer.cpp | 36 ++++++++++++++++++++++ 2 files changed, 60 insertions(+), 2 deletions(-) diff --git a/src/particles/ImpactXParticleContainer.H b/src/particles/ImpactXParticleContainer.H index 73c321207..f2c8c0a0f 100644 --- a/src/particles/ImpactXParticleContainer.H +++ b/src/particles/ImpactXParticleContainer.H @@ -72,6 +72,28 @@ namespace impactx }; }; + class ParIter + : public amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs> + { + public: + using amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>::ParIter; + + ParIter (ContainerType& pc, int level); + + ParIter (ContainerType& pc, int level, amrex::MFItInfo& info); + }; + + class ParConstIter + : public amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs> + { + public: + using amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>::ParConstIter; + + ParConstIter (ContainerType& pc, int level); + + ParConstIter (ContainerType& pc, int level, amrex::MFItInfo& info); + }; + /** Beam Particles in ImpactX * * This class stores particles, distributed over MPI ranks. @@ -81,10 +103,10 @@ namespace impactx { public: //! amrex iterator for particle boxes - using iterator = amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>; + using iterator = impactx::ParIter; //! amrex constant iterator for particle boxes (read-only) - using const_iterator = amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>; + using const_iterator = impactx::ParConstIter; //! Construct a new particle container ImpactXParticleContainer (amrex::AmrCore* amr_core); diff --git a/src/particles/ImpactXParticleContainer.cpp b/src/particles/ImpactXParticleContainer.cpp index 864b59a1a..45b81aecc 100644 --- a/src/particles/ImpactXParticleContainer.cpp +++ b/src/particles/ImpactXParticleContainer.cpp @@ -24,6 +24,42 @@ namespace impactx { + ParIter::ParIter (ContainerType& pc, int level) + : amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, + amrex::MFItInfo().SetDynamic( []() -> bool { + bool do_dynamic = false; + amrex::ParmParse pp_impactx("impactx"); + pp_impactx.query("do_dynamic_scheduling", do_dynamic); + return do_dynamic; }())) + {} + + ParIter::ParIter (ContainerType& pc, int level, amrex::MFItInfo& info) + : amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, + info.SetDynamic( []() -> bool { + bool do_dynamic = false; + amrex::ParmParse pp_impactx("impactx"); + pp_impactx.query("do_dynamic_scheduling", do_dynamic); + return do_dynamic; }())) + {} + + ParConstIter::ParConstIter (ContainerType& pc, int level) + : amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, + amrex::MFItInfo().SetDynamic( []() -> bool { + bool do_dynamic = false; + amrex::ParmParse pp_impactx("impactx"); + pp_impactx.query("do_dynamic_scheduling", do_dynamic); + return do_dynamic; }())) + {} + + ParConstIter::ParConstIter (ContainerType& pc, int level, amrex::MFItInfo& info) + : amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, + info.SetDynamic( []() -> bool { + bool do_dynamic = false; + amrex::ParmParse pp_impactx("impactx"); + pp_impactx.query("do_dynamic_scheduling", do_dynamic); + return do_dynamic; }())) + {} + ImpactXParticleContainer::ImpactXParticleContainer (amrex::AmrCore* amr_core) : amrex::ParticleContainer<0, 0, RealSoA::nattribs, IntSoA::nattribs>(amr_core->GetParGDB()) { From eea6241357dee1aab1675b70e158356e96d048d0 Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Tue, 20 Sep 2022 16:45:00 -0700 Subject: [PATCH 4/8] do in nice way with named function --- src/particles/ImpactXParticleContainer.cpp | 35 +++++++--------------- 1 file changed, 11 insertions(+), 24 deletions(-) diff --git a/src/particles/ImpactXParticleContainer.cpp b/src/particles/ImpactXParticleContainer.cpp index 45b81aecc..d7a2f4e4d 100644 --- a/src/particles/ImpactXParticleContainer.cpp +++ b/src/particles/ImpactXParticleContainer.cpp @@ -24,41 +24,28 @@ namespace impactx { + bool do_omp_dynamic () { + bool do_dynamic = false; + amrex::ParmParse pp_impactx("impactx"); + pp_impactx.query("do_dynamic_scheduling", do_dynamic); + return do_dynamic; + } + ParIter::ParIter (ContainerType& pc, int level) : amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, - amrex::MFItInfo().SetDynamic( []() -> bool { - bool do_dynamic = false; - amrex::ParmParse pp_impactx("impactx"); - pp_impactx.query("do_dynamic_scheduling", do_dynamic); - return do_dynamic; }())) - {} + amrex::MFItInfo().SetDynamic(do_omp_dynamic())) {} ParIter::ParIter (ContainerType& pc, int level, amrex::MFItInfo& info) : amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, - info.SetDynamic( []() -> bool { - bool do_dynamic = false; - amrex::ParmParse pp_impactx("impactx"); - pp_impactx.query("do_dynamic_scheduling", do_dynamic); - return do_dynamic; }())) - {} + info.SetDynamic(do_omp_dynamic())) {} ParConstIter::ParConstIter (ContainerType& pc, int level) : amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, - amrex::MFItInfo().SetDynamic( []() -> bool { - bool do_dynamic = false; - amrex::ParmParse pp_impactx("impactx"); - pp_impactx.query("do_dynamic_scheduling", do_dynamic); - return do_dynamic; }())) - {} + amrex::MFItInfo().SetDynamic(do_omp_dynamic())) {} ParConstIter::ParConstIter (ContainerType& pc, int level, amrex::MFItInfo& info) : amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs>(pc, level, - info.SetDynamic( []() -> bool { - bool do_dynamic = false; - amrex::ParmParse pp_impactx("impactx"); - pp_impactx.query("do_dynamic_scheduling", do_dynamic); - return do_dynamic; }())) - {} + info.SetDynamic(do_omp_dynamic())) {} ImpactXParticleContainer::ImpactXParticleContainer (amrex::AmrCore* amr_core) : amrex::ParticleContainer<0, 0, RealSoA::nattribs, IntSoA::nattribs>(amr_core->GetParGDB()) From 6e0c5946ef12b5d493918933cc5376548df37855 Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Wed, 21 Sep 2022 10:03:19 -0700 Subject: [PATCH 5/8] add some docstrings --- src/particles/ImpactXParticleContainer.H | 10 ++++++++++ 1 file changed, 10 insertions(+) diff --git a/src/particles/ImpactXParticleContainer.H b/src/particles/ImpactXParticleContainer.H index f2c8c0a0f..6744cf2fd 100644 --- a/src/particles/ImpactXParticleContainer.H +++ b/src/particles/ImpactXParticleContainer.H @@ -72,6 +72,11 @@ namespace impactx }; }; + /** AMReX iterator for particle boxes + * + * We subclass here to change the default threading strategy, which is + * `static` in AMReX, to `dynamic` in ImpactX. + */ class ParIter : public amrex::ParIter<0, 0, RealSoA::nattribs, IntSoA::nattribs> { @@ -83,6 +88,11 @@ namespace impactx ParIter (ContainerType& pc, int level, amrex::MFItInfo& info); }; + /** Const AMReX iterator for particle boxes - data is read only. + * + * We subclass here to change the default threading strategy, which is + * `static` in AMReX, to `dynamic` in ImpactX. + */ class ParConstIter : public amrex::ParConstIter<0, 0, RealSoA::nattribs, IntSoA::nattribs> { From 91419a6c053940382c5d46ebc69b3c165f278787 Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Wed, 21 Sep 2022 10:14:51 -0700 Subject: [PATCH 6/8] also enable omp in the new spacecharge particle gather and push --- src/particles/spacecharge/GatherAndPush.cpp | 3 +++ 1 file changed, 3 insertions(+) diff --git a/src/particles/spacecharge/GatherAndPush.cpp b/src/particles/spacecharge/GatherAndPush.cpp index 74d7111c4..edd3343e6 100644 --- a/src/particles/spacecharge/GatherAndPush.cpp +++ b/src/particles/spacecharge/GatherAndPush.cpp @@ -43,6 +43,9 @@ namespace impactx::spacecharge // loop over all particle boxes using ParIt = ImpactXParticleContainer::iterator; +#ifdef AMREX_USE_OMP +#pragma omp parallel if (amrex::Gpu::notInLaunchRegion()) +#endif for (ParIt pti(pc, lev); pti.isValid(); ++pti) { const int np = pti.numParticles(); From 49d3f6f811a735380d55f18791f271dd6b1834b6 Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Wed, 21 Sep 2022 10:24:12 -0700 Subject: [PATCH 7/8] turn on dynamic scheduling by default --- src/particles/ImpactXParticleContainer.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/src/particles/ImpactXParticleContainer.cpp b/src/particles/ImpactXParticleContainer.cpp index d7a2f4e4d..9f80c3955 100644 --- a/src/particles/ImpactXParticleContainer.cpp +++ b/src/particles/ImpactXParticleContainer.cpp @@ -25,7 +25,7 @@ namespace impactx { bool do_omp_dynamic () { - bool do_dynamic = false; + bool do_dynamic = true; amrex::ParmParse pp_impactx("impactx"); pp_impactx.query("do_dynamic_scheduling", do_dynamic); return do_dynamic; From 10d5e87cdcf62715fb3cc5005c97422aa5962fe3 Mon Sep 17 00:00:00 2001 From: Andrew Myers Date: Thu, 22 Sep 2022 11:58:37 -0700 Subject: [PATCH 8/8] also accelerate mfiter loop with omp --- src/particles/spacecharge/ForceFromSelfFields.cpp | 3 +++ 1 file changed, 3 insertions(+) diff --git a/src/particles/spacecharge/ForceFromSelfFields.cpp b/src/particles/spacecharge/ForceFromSelfFields.cpp index d5287c253..d92be8cd7 100644 --- a/src/particles/spacecharge/ForceFromSelfFields.cpp +++ b/src/particles/spacecharge/ForceFromSelfFields.cpp @@ -40,6 +40,9 @@ namespace impactx::spacecharge space_charge_field.at(lev).at("y").setVal(0.); space_charge_field.at(lev).at("z").setVal(0.); +#ifdef AMREX_USE_OMP +#pragma omp parallel if (amrex::Gpu::notInLaunchRegion()) +#endif for (amrex::MFIter mfi(phi.at(lev)); mfi.isValid(); ++mfi) { amrex::Box bx = mfi.validbox();