diff options
| author |  Remi Lehe <remi.lehe@normalesup.org>
| 2019-09-25 14:53:21 -0700 |
|---|---|---|
| committer | Remi Lehe <remi.lehe@normalesup.org>
| 2019-10-01 16:32:38 -0700 |
| commit | feac8d56e5cc84f63f6e7be7e1668113558c9e47 (patch) | |
| tree | 22535973a3daf9df09ccbc68adc17f3e7504d979 /Source/Particles/Sorting | |
| parent | a7d4ebb9cb2b7528b7b1891eccb794fb6b70cf09 (diff) | |
| download | WarpX-feac8d56e5cc84f63f6e7be7e1668113558c9e47.tar.gz WarpX-feac8d56e5cc84f63f6e7be7e1668113558c9e47.tar.zst WarpX-feac8d56e5cc84f63f6e7be7e1668113558c9e47.zip | |
Implement `iota` on GPU
Diffstat (limited to 'Source/Particles/Sorting')
| -rw-r--r-- | Source/Particles/Sorting/Make.package | 1 | ||||
| -rw-r--r-- | Source/Particles/Sorting/Partition.cpp | 10 | ||||
| -rw-r--r-- | Source/Particles/Sorting/SortingUtils.H | 18 |
3 files changed, 25 insertions, 4 deletions
diff --git a/Source/Particles/Sorting/Make.package b/Source/Particles/Sorting/Make.package index 750d2607e..f9c708e5b 100644 --- a/Source/Particles/Sorting/Make.package +++ b/Source/Particles/Sorting/Make.package @@ -1,3 +1,4 @@ +CEXE_headers += SortingUtils.H CEXE_sources += Partition.cpp INCLUDE_LOCATIONS += $(WARPX_HOME)/Source/Particles/Sorting VPATH_LOCATIONS += $(WARPX_HOME)/Source/Particles/Sorting diff --git a/Source/Particles/Sorting/Partition.cpp b/Source/Particles/Sorting/Partition.cpp index 29e187821..49d4d25a6 100644 --- a/Source/Particles/Sorting/Partition.cpp +++ b/Source/Particles/Sorting/Partition.cpp @@ -1,5 +1,6 @@ -#include <WarpX.H> +#include <SortingUtils.H> #include <PhysicalParticleContainer.H> +#include <WarpX.H> #include <AMReX_Particles.H> using namespace amrex; @@ -43,7 +44,7 @@ PhysicalParticleContainer::PartitionParticlesInBuffers( Gpu::ManagedDeviceVector<int> inexflag; inexflag.resize(np); - // We need to partition the large buffer first + // Select the largest buffer first iMultiFab const* bmasks = (WarpX::n_field_gather_buffer >= WarpX::n_current_deposition_buffer) ? gather_masks : current_masks; @@ -60,9 +61,10 @@ PhysicalParticleContainer::PartitionParticlesInBuffers( } ); - Vector<long> pid; + // Partition the particles according whether they are in the large buffer or not + Gpu::ManagedDeviceVector<long> pid; pid.resize(np); - std::iota(pid.begin(), pid.end(), 0L); + fillWithConsecutiveIntegers( pid ); auto sep = std::stable_partition(pid.begin(), pid.end(), [&inexflag](long id) { return inexflag[id]; }); diff --git a/Source/Particles/Sorting/SortingUtils.H b/Source/Particles/Sorting/SortingUtils.H new file mode 100644 index 000000000..d072ebd2a --- /dev/null +++ b/Source/Particles/Sorting/SortingUtils.H @@ -0,0 +1,18 @@ +#ifndef WARPX_PARTICLES_SORTING_SORTINGUTILS_H_ +#define WARPX_PARTICLES_SORTING_SORTINGUTILS_H_ + +#include <AMReX_Gpu.H> +#include <AMReX_CudaContainers.H> + +// TODO: Add documentation +void fillWithConsecutiveIntegers( amrex::Gpu::ManagedDeviceVector<long>& v ) { +#ifdef AMREX_USE_GPU + // On GPU: Use thrust + thrust::sequence( v.begin(), v.end() ); +#else + // On CPU: Use std library + std::iota( v.begin(), v.end(), 0L ); +#endif +} + +#endif // WARPX_PARTICLES_SORTING_SORTINGUTILS_H_ |