10
10
// File created by: Jeongnim Kim, [email protected] , Intel Corp.
11
11
// ////////////////////////////////////////////////////////////////////////////////////
12
12
// -*- C++ -*-
13
- #ifndef QMCPLUSPLUS_DTDIMPL_AB_OMP_H
14
- #define QMCPLUSPLUS_DTDIMPL_AB_OMP_H
13
+ #ifndef QMCPLUSPLUS_DTDIMPL_AB_OMPTARGET_H
14
+ #define QMCPLUSPLUS_DTDIMPL_AB_OMPTARGET_H
15
15
16
- #include " OpenMP /OMPallocator.hpp"
16
+ #include " OMPTarget /OMPallocator.hpp"
17
17
#include " Platforms/PinnedAllocator.h"
18
- #include " Particle/RealSpacePositionsOMP .h"
18
+ #include " Particle/RealSpacePositionsOMPTarget .h"
19
19
20
20
namespace qmcplusplus
21
21
{
22
22
/* *@ingroup nnlist
23
23
* @brief A derived classe from DistacneTableData, specialized for AB using a transposed form
24
24
*/
25
25
template <typename T, unsigned D, int SC>
26
- class SoaDistanceTableABOMP : public DTD_BConds <T, D, SC>, public DistanceTableData
26
+ class SoaDistanceTableABOMPTarget : public DTD_BConds <T, D, SC>, public DistanceTableData
27
27
{
28
28
private:
29
29
template <typename DT>
@@ -49,15 +49,15 @@ class SoaDistanceTableABOMP : public DTD_BConds<T, D, SC>, public DistanceTableD
49
49
NewTimer& eval_timer_;
50
50
51
51
public:
52
- SoaDistanceTableABOMP (const ParticleSet& source, ParticleSet& target)
52
+ SoaDistanceTableABOMPTarget (const ParticleSet& source, ParticleSet& target)
53
53
: DTD_BConds<T, D, SC>(source.Lattice),
54
54
DistanceTableData (source, target),
55
55
r_dr_device_ptr_ (nullptr ),
56
- offload_timer_ (*timer_manager.createTimer(std::string(" SoaDistanceTableABOMP ::offload_" ) + target.getName() + " _" + source.getName(), timer_level_fine)),
57
- copy_timer_ (*timer_manager.createTimer(std::string(" SoaDistanceTableABOMP ::copy_" ) + target.getName() + " _" + source.getName(), timer_level_fine)),
58
- eval_timer_ (*timer_manager.createTimer(std::string(" SoaDistanceTableABOMP ::evaluate_" ) + target.getName() + " _" + source.getName(), timer_level_fine))
56
+ offload_timer_ (*timer_manager.createTimer(std::string(" SoaDistanceTableABOMPTarget ::offload_" ) + target.getName() + " _" + source.getName(), timer_level_fine)),
57
+ copy_timer_ (*timer_manager.createTimer(std::string(" SoaDistanceTableABOMPTarget ::copy_" ) + target.getName() + " _" + source.getName(), timer_level_fine)),
58
+ eval_timer_ (*timer_manager.createTimer(std::string(" SoaDistanceTableABOMPTarget ::evaluate_" ) + target.getName() + " _" + source.getName(), timer_level_fine))
59
59
{
60
- auto * coordinates_soa = dynamic_cast <const RealSpacePositionsOMP *>(&source.getCoordinates ());
60
+ auto * coordinates_soa = dynamic_cast <const RealSpacePositionsOMPTarget *>(&source.getCoordinates ());
61
61
if (!coordinates_soa) throw std::runtime_error (" Source particle set doesn't have OpenMP offload. Contact developers!" );
62
62
resize (source.getTotalNum (), target.getTotalNum ());
63
63
#pragma omp target enter data map(to:this[:1])
@@ -94,10 +94,10 @@ class SoaDistanceTableABOMP : public DTD_BConds<T, D, SC>, public DistanceTableD
94
94
temp_dr_.resize (N_sources);
95
95
}
96
96
97
- SoaDistanceTableABOMP () = delete ;
98
- SoaDistanceTableABOMP (const SoaDistanceTableABOMP &) = delete ;
97
+ SoaDistanceTableABOMPTarget () = delete ;
98
+ SoaDistanceTableABOMPTarget (const SoaDistanceTableABOMPTarget &) = delete ;
99
99
100
- ~SoaDistanceTableABOMP ()
100
+ ~SoaDistanceTableABOMPTarget ()
101
101
{
102
102
#pragma omp target exit data map(delete:this[:1])
103
103
}
@@ -187,13 +187,13 @@ class SoaDistanceTableABOMP : public DTD_BConds<T, D, SC>, public DistanceTableD
187
187
count_targets = 0 ;
188
188
for (size_t iw = 0 ; iw < nw; iw++)
189
189
{
190
- auto & dt = static_cast <SoaDistanceTableABOMP &>(dt_list[iw].get ());
190
+ auto & dt = static_cast <SoaDistanceTableABOMPTarget &>(dt_list[iw].get ());
191
191
ParticleSet& pset (p_list[iw]);
192
192
193
193
assert (N_sources == dt.N_sources );
194
194
195
- auto & RSoA_OMP = static_cast <const RealSpacePositionsOMP &>(dt.Origin ->getCoordinates ());
196
- source_ptrs[iw] = const_cast <RealType*>(RSoA_OMP .getDevicePtr ());
195
+ auto & RSoA_OMPTarget = static_cast <const RealSpacePositionsOMPTarget &>(dt.Origin ->getCoordinates ());
196
+ source_ptrs[iw] = const_cast <RealType*>(RSoA_OMPTarget .getDevicePtr ());
197
197
198
198
for (size_t iat = 0 ; iat < pset.getTotalNum (); ++iat, ++count_targets)
199
199
{
@@ -242,7 +242,7 @@ class SoaDistanceTableABOMP : public DTD_BConds<T, D, SC>, public DistanceTableD
242
242
ScopedTimer copy (©_timer_);
243
243
for (size_t iw = 0 ; iw < nw; iw++)
244
244
{
245
- auto & dt = static_cast <SoaDistanceTableABOMP &>(dt_list[iw].get ());
245
+ auto & dt = static_cast <SoaDistanceTableABOMPTarget &>(dt_list[iw].get ());
246
246
auto * pool_ptr = dt.r_dr_memorypool_ .data ();
247
247
#pragma omp target update from(pool_ptr[:dt.r_dr_memorypool_.size()]) nowait depend(inout:total_targets)
248
248
}
@@ -279,13 +279,13 @@ class SoaDistanceTableABOMP : public DTD_BConds<T, D, SC>, public DistanceTableD
279
279
count_targets = 0 ;
280
280
for (size_t iw = 0 ; iw < nw; iw++)
281
281
{
282
- auto & dt = static_cast <SoaDistanceTableABOMP &>(dt_list[iw].get ());
282
+ auto & dt = static_cast <SoaDistanceTableABOMPTarget &>(dt_list[iw].get ());
283
283
ParticleSet& pset (p_list[iw]);
284
284
285
285
assert (N_sources == dt.N_sources );
286
286
287
- auto & RSoA_OMP = static_cast <const RealSpacePositionsOMP &>(dt.Origin ->getCoordinates ());
288
- source_ptrs[iw] = const_cast <RealType*>(RSoA_OMP .getDevicePtr ());
287
+ auto & RSoA_OMPTarget = static_cast <const RealSpacePositionsOMPTarget &>(dt.Origin ->getCoordinates ());
288
+ source_ptrs[iw] = const_cast <RealType*>(RSoA_OMPTarget .getDevicePtr ());
289
289
290
290
for (size_t iat = 0 ; iat < pset.getTotalNum (); ++iat, ++count_targets)
291
291
{
@@ -337,7 +337,7 @@ class SoaDistanceTableABOMP : public DTD_BConds<T, D, SC>, public DistanceTableD
337
337
{
338
338
const int wid = walker_id_ptr[iat];
339
339
const int pid = particle_id[iat];
340
- auto & dt = static_cast <SoaDistanceTableABOMP &>(dt_list[wid].get ());
340
+ auto & dt = static_cast <SoaDistanceTableABOMPTarget &>(dt_list[wid].get ());
341
341
assert (N_sources_padded == dt.displacements_ [pid].capacity ());
342
342
auto offset = offload_output.data () + iat * N_sources_padded * (D + 1 );
343
343
std::copy_n (offset, N_sources_padded, dt.distances_ [pid].data ());
0 commit comments