diff options
Diffstat (limited to 'Source/Particles/WarpXParticleContainer.cpp')
| -rw-r--r-- | Source/Particles/WarpXParticleContainer.cpp | 24 |
1 files changed, 12 insertions, 12 deletions
diff --git a/Source/Particles/WarpXParticleContainer.cpp b/Source/Particles/WarpXParticleContainer.cpp index c98bc7098..65c7baeaa 100644 --- a/Source/Particles/WarpXParticleContainer.cpp +++ b/Source/Particles/WarpXParticleContainer.cpp @@ -1,7 +1,7 @@ /* Copyright 2019-2020 Andrew Myers, Axel Huebl, David Grote * Jean-Luc Vay, Luca Fedeli, Maxence Thevenet - * Remi Lehe, Revathi Jambunathan, Weiqun Zhang - * Yinjian Zhao, levinem + * Michael Rowan, Remi Lehe, Revathi Jambunathan + * Weiqun Zhang, Yinjian Zhao, levinem * * This file is part of WarpX. * @@ -728,7 +728,7 @@ WarpXParticleContainer::PushX (int lev, amrex::Real dt) if (do_not_push) return; - MultiFab* cost = WarpX::getCosts(lev); + amrex::Vector<amrex::Real>* cost = WarpX::getCosts(lev); #ifdef _OPENMP #pragma omp parallel @@ -737,6 +737,10 @@ WarpXParticleContainer::PushX (int lev, amrex::Real dt) for (WarpXParIter pti(*this, lev); pti.isValid(); ++pti) { + if (cost && WarpX::load_balance_costs_update_algo == LoadBalanceCostsUpdateAlgo::Timers) + { + amrex::Gpu::synchronize(); + } Real wt = amrex::second(); // @@ -766,15 +770,11 @@ WarpXParticleContainer::PushX (int lev, amrex::Real dt) } ); - if (cost) { - const Box& tbx = pti.tilebox(); - wt = (amrex::second() - wt) / tbx.d_numPts(); - Array4<Real> const& costarr = cost->array(pti); - amrex::ParallelFor(tbx, - [=] AMREX_GPU_DEVICE (int i, int j, int k) noexcept - { - costarr(i,j,k) += wt; - }); + if (cost && WarpX::load_balance_costs_update_algo == LoadBalanceCostsUpdateAlgo::Timers) + { + amrex::Gpu::synchronize(); + wt = amrex::second() - wt; + amrex::HostDevice::Atomic::Add( &(*cost)[pti.index()], wt); } } } |