From 66f370706bc3133f267406b1a7d3f2f046051ea2 Mon Sep 17 00:00:00 2001 From: MaxThevenet Date: Mon, 11 Nov 2019 13:24:32 -0800 Subject: Call RedistributeGPU on GPU, since @amtyers fixed the issue with runtime components --- Source/Particles/PhysicalParticleContainer.cpp | 10 ---------- 1 file changed, 10 deletions(-) (limited to 'Source/Particles/PhysicalParticleContainer.cpp') diff --git a/Source/Particles/PhysicalParticleContainer.cpp b/Source/Particles/PhysicalParticleContainer.cpp index 51690d659..77f6601c6 100644 --- a/Source/Particles/PhysicalParticleContainer.cpp +++ b/Source/Particles/PhysicalParticleContainer.cpp @@ -62,16 +62,6 @@ PhysicalParticleContainer::PhysicalParticleContainer (AmrCore* amr_core, int isp "Radiation reaction can be enabled only if Boris pusher is used"); //_____________________________ -#ifdef AMREX_USE_GPU - Print()<<"\n-----------------------------------------------------\n"; - Print()<<"WARNING: field ionization on GPU uses RedistributeCPU\n"; - Print()<<"-----------------------------------------------------\n\n"; - //AMREX_ALWAYS_ASSERT_WITH_MESSAGE( - //do_field_ionization == 0, - //"Field ionization does not work on GPU so far, because the current " - //"version of Redistribute in AMReX does not work with runtime parameters"); -#endif - #ifdef WARPX_QED //Add real component if QED is enabled pp.query("do_qed", m_do_qed); -- cgit v1.2.3 From 25c55bf3ecd7c6ca2d799fd55244e1a8b5f70970 Mon Sep 17 00:00:00 2001 From: Weiqun Zhang Date: Mon, 11 Nov 2019 21:10:12 -0800 Subject: Cuda -> Gpu according to AMReX change --- Source/Diagnostics/BackTransformedDiagnostic.H | 1 - Source/Filter/Filter.H | 1 - Source/Laser/LaserParticleContainer.cpp | 2 +- Source/Particles/PhotonParticleContainer.H | 6 +++--- Source/Particles/PhotonParticleContainer.cpp | 6 +++--- Source/Particles/PhysicalParticleContainer.H | 6 +++--- Source/Particles/PhysicalParticleContainer.cpp | 8 ++++---- Source/Particles/RigidInjectedParticleContainer.H | 6 +++--- Source/Particles/RigidInjectedParticleContainer.cpp | 12 ++++++------ Source/Particles/Sorting/SortingUtils.H | 3 +-- Source/Particles/WarpXParticleContainer.H | 12 ++++++------ 11 files changed, 30 insertions(+), 33 deletions(-) (limited to 'Source/Particles/PhysicalParticleContainer.cpp') diff --git a/Source/Diagnostics/BackTransformedDiagnostic.H b/Source/Diagnostics/BackTransformedDiagnostic.H index 9e24caa1b..94d557605 100644 --- a/Source/Diagnostics/BackTransformedDiagnostic.H +++ b/Source/Diagnostics/BackTransformedDiagnostic.H @@ -8,7 +8,6 @@ #include #include #include -#include #include "MultiParticleContainer.H" #include "WarpXConst.H" diff --git a/Source/Filter/Filter.H b/Source/Filter/Filter.H index 5eb84b96a..1ecb1bff4 100644 --- a/Source/Filter/Filter.H +++ b/Source/Filter/Filter.H @@ -1,5 +1,4 @@ #include -#include #ifndef WARPX_FILTER_H_ #define WARPX_FILTER_H_ diff --git a/Source/Laser/LaserParticleContainer.cpp b/Source/Laser/LaserParticleContainer.cpp index 9493672e0..35eadf064 100644 --- a/Source/Laser/LaserParticleContainer.cpp +++ b/Source/Laser/LaserParticleContainer.cpp @@ -454,7 +454,7 @@ LaserParticleContainer::Evolve (int lev, int const thread_num = 0; #endif - Cuda::ManagedDeviceVector plane_Xp, plane_Yp, amplitude_E; + Gpu::ManagedDeviceVector plane_Xp, plane_Yp, amplitude_E; for (WarpXParIter pti(*this, lev); pti.isValid(); ++pti) { diff --git a/Source/Particles/PhotonParticleContainer.H b/Source/Particles/PhotonParticleContainer.H index 9132cf4a9..851726d57 100644 --- a/Source/Particles/PhotonParticleContainer.H +++ b/Source/Particles/PhotonParticleContainer.H @@ -41,9 +41,9 @@ public: DtType a_dt_type=DtType::Full) override; virtual void PushPX(WarpXParIter& pti, - amrex::Cuda::ManagedDeviceVector& xp, - amrex::Cuda::ManagedDeviceVector& yp, - amrex::Cuda::ManagedDeviceVector& zp, + amrex::Gpu::ManagedDeviceVector& xp, + amrex::Gpu::ManagedDeviceVector& yp, + amrex::Gpu::ManagedDeviceVector& zp, amrex::Real dt, DtType a_dt_type=DtType::Full) override; // Don't push momenta for photons diff --git a/Source/Particles/PhotonParticleContainer.cpp b/Source/Particles/PhotonParticleContainer.cpp index 7e52b52e1..fd47ac8a0 100644 --- a/Source/Particles/PhotonParticleContainer.cpp +++ b/Source/Particles/PhotonParticleContainer.cpp @@ -51,9 +51,9 @@ void PhotonParticleContainer::InitData() void PhotonParticleContainer::PushPX(WarpXParIter& pti, - Cuda::ManagedDeviceVector& xp, - Cuda::ManagedDeviceVector& yp, - Cuda::ManagedDeviceVector& zp, + Gpu::ManagedDeviceVector& xp, + Gpu::ManagedDeviceVector& yp, + Gpu::ManagedDeviceVector& zp, Real dt, DtType a_dt_type) { diff --git a/Source/Particles/PhysicalParticleContainer.H b/Source/Particles/PhysicalParticleContainer.H index 17a504719..174488eb9 100644 --- a/Source/Particles/PhysicalParticleContainer.H +++ b/Source/Particles/PhysicalParticleContainer.H @@ -94,9 +94,9 @@ public: DtType a_dt_type=DtType::Full) override; virtual void PushPX(WarpXParIter& pti, - amrex::Cuda::ManagedDeviceVector& xp, - amrex::Cuda::ManagedDeviceVector& yp, - amrex::Cuda::ManagedDeviceVector& zp, + amrex::Gpu::ManagedDeviceVector& xp, + amrex::Gpu::ManagedDeviceVector& yp, + amrex::Gpu::ManagedDeviceVector& zp, amrex::Real dt, DtType a_dt_type=DtType::Full); virtual void PushP (int lev, amrex::Real dt, diff --git a/Source/Particles/PhysicalParticleContainer.cpp b/Source/Particles/PhysicalParticleContainer.cpp index 77f6601c6..938c80de5 100644 --- a/Source/Particles/PhysicalParticleContainer.cpp +++ b/Source/Particles/PhysicalParticleContainer.cpp @@ -1396,7 +1396,7 @@ PhysicalParticleContainer::SplitParticles(int lev) { auto& mypc = WarpX::GetInstance().GetPartContainer(); auto& pctmp_split = mypc.GetPCtmp(); - Cuda::ManagedDeviceVector xp, yp, zp; + Gpu::ManagedDeviceVector xp, yp, zp; RealVector psplit_x, psplit_y, psplit_z, psplit_w; RealVector psplit_ux, psplit_uy, psplit_uz; long np_split_to_add = 0; @@ -1554,9 +1554,9 @@ PhysicalParticleContainer::SplitParticles(int lev) void PhysicalParticleContainer::PushPX(WarpXParIter& pti, - Cuda::ManagedDeviceVector& xp, - Cuda::ManagedDeviceVector& yp, - Cuda::ManagedDeviceVector& zp, + Gpu::ManagedDeviceVector& xp, + Gpu::ManagedDeviceVector& yp, + Gpu::ManagedDeviceVector& zp, Real dt, DtType a_dt_type) { diff --git a/Source/Particles/RigidInjectedParticleContainer.H b/Source/Particles/RigidInjectedParticleContainer.H index 3abbb4afe..a2473c5ad 100644 --- a/Source/Particles/RigidInjectedParticleContainer.H +++ b/Source/Particles/RigidInjectedParticleContainer.H @@ -44,9 +44,9 @@ public: DtType a_dt_type=DtType::Full) override; virtual void PushPX(WarpXParIter& pti, - amrex::Cuda::ManagedDeviceVector& xp, - amrex::Cuda::ManagedDeviceVector& yp, - amrex::Cuda::ManagedDeviceVector& zp, + amrex::Gpu::ManagedDeviceVector& xp, + amrex::Gpu::ManagedDeviceVector& yp, + amrex::Gpu::ManagedDeviceVector& zp, amrex::Real dt, DtType a_dt_type=DtType::Full) override; virtual void PushP (int lev, amrex::Real dt, diff --git a/Source/Particles/RigidInjectedParticleContainer.cpp b/Source/Particles/RigidInjectedParticleContainer.cpp index 507206041..bee71fba1 100644 --- a/Source/Particles/RigidInjectedParticleContainer.cpp +++ b/Source/Particles/RigidInjectedParticleContainer.cpp @@ -78,7 +78,7 @@ RigidInjectedParticleContainer::RemapParticles() // Note that the particles are already in the boosted frame. // This value is saved to advance the particles not injected yet - Cuda::ManagedDeviceVector xp, yp, zp; + Gpu::ManagedDeviceVector xp, yp, zp; for (WarpXParIter pti(*this, lev); pti.isValid(); ++pti) { @@ -138,7 +138,7 @@ RigidInjectedParticleContainer::BoostandRemapParticles() #pragma omp parallel #endif { - Cuda::ManagedDeviceVector xp, yp, zp; + Gpu::ManagedDeviceVector xp, yp, zp; for (WarpXParIter pti(*this, 0); pti.isValid(); ++pti) { @@ -209,9 +209,9 @@ RigidInjectedParticleContainer::BoostandRemapParticles() void RigidInjectedParticleContainer::PushPX(WarpXParIter& pti, - Cuda::ManagedDeviceVector& xp, - Cuda::ManagedDeviceVector& yp, - Cuda::ManagedDeviceVector& zp, + Gpu::ManagedDeviceVector& xp, + Gpu::ManagedDeviceVector& yp, + Gpu::ManagedDeviceVector& zp, Real dt, DtType a_dt_type) { @@ -222,7 +222,7 @@ RigidInjectedParticleContainer::PushPX(WarpXParIter& pti, auto& uzp = attribs[PIdx::uz]; // Save the position and momenta, making copies - Cuda::ManagedDeviceVector xp_save, yp_save, zp_save; + Gpu::ManagedDeviceVector xp_save, yp_save, zp_save; RealVector uxp_save, uyp_save, uzp_save; ParticleReal* const AMREX_RESTRICT x = xp.dataPtr(); diff --git a/Source/Particles/Sorting/SortingUtils.H b/Source/Particles/Sorting/SortingUtils.H index 28a1b73b7..80eaaf9cb 100644 --- a/Source/Particles/Sorting/SortingUtils.H +++ b/Source/Particles/Sorting/SortingUtils.H @@ -2,7 +2,6 @@ #define WARPX_PARTICLES_SORTING_SORTINGUTILS_H_ #include -#include #include #ifdef AMREX_USE_GPU #include @@ -43,7 +42,7 @@ ForwardIterator stablePartition(ForwardIterator const index_begin, // On GPU: Use thrust int const* AMREX_RESTRICT predicate_ptr = predicate.dataPtr(); ForwardIterator const sep = thrust::stable_partition( - thrust::cuda::par(amrex::Cuda::The_ThrustCachedAllocator()), + thrust::cuda::par(amrex::Gpu::The_ThrustCachedAllocator()), index_begin, index_end, [predicate_ptr] AMREX_GPU_DEVICE (long i) { return predicate_ptr[i]; } ); diff --git a/Source/Particles/WarpXParticleContainer.H b/Source/Particles/WarpXParticleContainer.H index dbd913c5b..7f86930b2 100644 --- a/Source/Particles/WarpXParticleContainer.H +++ b/Source/Particles/WarpXParticleContainer.H @@ -73,12 +73,12 @@ public: WarpXParIter (ContainerType& pc, int level); #if (AMREX_SPACEDIM == 2) - void GetPosition (amrex::Cuda::ManagedDeviceVector& x, - amrex::Cuda::ManagedDeviceVector& y, - amrex::Cuda::ManagedDeviceVector& z) const; - void SetPosition (const amrex::Cuda::ManagedDeviceVector& x, - const amrex::Cuda::ManagedDeviceVector& y, - const amrex::Cuda::ManagedDeviceVector& z); + void GetPosition (amrex::Gpu::ManagedDeviceVector& x, + amrex::Gpu::ManagedDeviceVector& y, + amrex::Gpu::ManagedDeviceVector& z) const; + void SetPosition (const amrex::Gpu::ManagedDeviceVector& x, + const amrex::Gpu::ManagedDeviceVector& y, + const amrex::Gpu::ManagedDeviceVector& z); #endif const std::array& GetAttribs () const { return GetStructOfArrays().GetRealData(); -- cgit v1.2.3