OpenFPM_pdata  4.1.0
Project that contain the implementation of distributed structures
memory_conf.hpp
1 /*
2  * memory_conf.hpp
3  *
4  * Created on: Aug 28, 2014
5  * Author: Pietro Incardona
6  */
7 
8 #ifndef MEMORY_CONF_HPP_
9 #define MEMORY_CONF_HPP_
10 
11 #include "util/variadic_to_vmpl.hpp"
12 #include "t_to_memory_c.hpp"
13 #include "Vector/vect_isel.hpp"
14 #include "Vector/util.hpp"
15 #include "util/tokernel_transformation.hpp"
16 #include "util/hostDevice_util_funcs.hpp"
17 
18 constexpr int SOA_layout_IA = 2;
19 constexpr int SOA_layout = 1;
20 constexpr int AOS_layout = 0;
21 
39 template<typename Seq>
40 struct inter_memc
41 {
42  typedef typename v_transform<t_to_memory_c,Seq>::type type;
43 };
44 
62 template<typename Seq>
64 {
65  typedef typename v_transform<t_to_memory_c_red,Seq>::type type;
66 };
67 
68 
82 template<typename T>
84 {
86  typedef typename inter_memc<typename T::type>::type type;
87 
89  typedef int yes_is_inte;
90 
91  typedef boost::mpl::int_<SOA_layout_IA> type_value;
92 
102  template<unsigned int p, typename data_type, typename g1_type, typename key_type>
103  __host__ __device__ static inline auto get(data_type & data_, const g1_type & g1, const key_type & v1) -> decltype(boost::fusion::at_c<p>(data_).mem_r.operator[](g1.LinId(v1)))
104  {
105  return boost::fusion::at_c<p>(data_).mem_r.operator[](g1.LinId(v1));
106  }
107 
117  template<unsigned int p, typename data_type, typename g1_type>
118  __host__ __device__ static inline auto get_lin(data_type & data_, const g1_type & g1, size_t lin_id) -> decltype(boost::fusion::at_c<p>(data_).mem_r.operator[](lin_id))
119  {
120  return boost::fusion::at_c<p>(data_).mem_r.operator[](lin_id);
121  }
122 
132  template<unsigned int p, typename data_type, typename g1_type, typename key_type>
133  __host__ __device__ static inline auto get_c(const data_type & data_, const g1_type & g1, const key_type & v1) -> decltype(boost::fusion::at_c<p>(data_).mem_r.operator[](g1.LinId(v1)))
134  {
135  return boost::fusion::at_c<p>(data_).mem_r.operator[](g1.LinId(v1));
136  }
137 
147  template<unsigned int p, typename data_type, typename g1_type>
148  __host__ __device__ static inline auto get_lin_c(const data_type & data_, const g1_type & g1, size_t lin_id) -> decltype(boost::fusion::at_c<p>(data_).mem_r.operator[](lin_id))
149  {
150  return boost::fusion::at_c<p>(data_).mem_r.operator[](lin_id);
151  }
152 
161  template<typename S, typename data_type, unsigned int ... prp>
162  static void hostToDevice(data_type & data_, size_t start, size_t stop)
163  {
164  host_to_device_impl<T,memory_traits_inte,S, prp ...> dth(data_,start,stop);
165 
166  boost::mpl::for_each_ref< boost::mpl::range_c<int,0,sizeof...(prp)> >(dth);
167  }
168 
177  template<typename data_type, unsigned int ... prp>
178  static void deviceToHost(data_type & data_, size_t start, size_t stop)
179  {
180  device_to_host_start_stop_impl<T, memory_traits_inte, prp ...> dth(data_,start,stop);
181 
182  boost::mpl::for_each_ref< boost::mpl::range_c<int,0,sizeof...(prp)> >(dth);
183  }
184 };
185 
199 /*template<typename T>
200 struct memory_traits_inte_red
201 {
203  typedef typename inter_memc_red<typename T::type>::type type;
204 
206  typedef int yes_is_inte;
207 };*/
208 
213 template<typename T, bool is_agg>
215 {
217 };
218 
223 template<typename T>
224 struct memory_traits_lin_type<T,false>
225 {
226  typedef void type;
227 };
228 
241 template<typename T>
243 {
246 
247  typedef int yes_is_tlin;
248 
249  typedef boost::mpl::int_<AOS_layout> type_value;
250 
260  template<unsigned int p, typename data_type, typename g1_type, typename key_type>
261  __host__ __device__ static inline auto get(data_type & data_, const g1_type & g1, const key_type & v1) -> decltype(boost::fusion::at_c<p>(data_.mem_r.operator[](g1.LinId(v1)))) &
262  {
263  return boost::fusion::at_c<p>(data_.mem_r.operator[](g1.LinId(v1)));
264  }
265 
275  template<unsigned int p, typename data_type, typename g1_type>
276  __host__ __device__ static inline auto get_lin(data_type & data_, const g1_type & g1, const size_t lin_id) -> decltype(boost::fusion::at_c<p>(data_.mem_r.operator[](lin_id))) &
277  {
278  return boost::fusion::at_c<p>(data_.mem_r.operator[](lin_id));
279  }
280 
290  template<unsigned int p, typename data_type, typename g1_type, typename key_type>
291  __host__ __device__ static inline auto get_c(const data_type & data_, const g1_type & g1, const key_type & v1) -> decltype(boost::fusion::at_c<p>(data_.mem_r.operator[](g1.LinId(v1)))) &
292  {
293  return boost::fusion::at_c<p>(data_.mem_r.operator[](g1.LinId(v1)));
294  }
295 
305  template<unsigned int p, typename data_type, typename g1_type>
306  __host__ __device__ static inline auto get_lin_c(const data_type & data_, const g1_type & g1, const size_t lin_id) -> decltype(boost::fusion::at_c<p>(data_.mem_r.operator[](lin_id))) &
307  {
308  return boost::fusion::at_c<p>(data_.mem_r.operator[](lin_id));
309  }
310 
319  template<typename S, typename data_type, unsigned int ... prp>
320  static void hostToDevice(data_type & data_, size_t start, size_t stop)
321  {
322  data_.mem->hostToDevice(start*sizeof(T),(stop+1)*sizeof(T));
323  }
324 
333  template<typename data_type, unsigned int ... prp>
334  static void deviceToHost(data_type & data_, size_t start, size_t stop)
335  {
336  data_.mem->deviceToHost(start*sizeof(T),(stop+1)*sizeof(T));
337  }
338 };
339 
340 
342 
343 template<typename T, typename Sfinae = void>
344 struct is_layout_mlin: std::false_type {};
345 
346 
356 template<typename T>
357 struct is_layout_mlin<T, typename Void< typename T::yes_is_tlin>::type> : std::true_type
358 {};
359 
360 
361 template<typename T, typename Sfinae = void>
362 struct is_layout_inte: std::false_type {};
363 
364 
374 template<typename T>
375 struct is_layout_inte<T, typename Void< typename T::yes_is_inte>::type> : std::true_type
376 {};
377 
389 template<typename T, unsigned int impl = is_vector<T>::value >
390 struct is_multiple_buffer_each_prp: std::false_type
391 {};
392 
393 template<typename T>
394 struct is_multiple_buffer_each_prp<T,true>: is_layout_inte<typename T::layout_base_>
395 {};
396 
397 #endif
__host__ static __device__ auto get_lin(data_type &data_, const g1_type &g1, const size_t lin_id) -> decltype(boost::fusion::at_c< p >(data_.mem_r.operator[](lin_id))) &
Return a reference to the selected element.
Transform the boost::fusion::vector into memory specification (memory_traits)
__host__ static __device__ auto get_lin_c(const data_type &data_, const g1_type &g1, size_t lin_id) -> decltype(boost::fusion::at_c< p >(data_).mem_r.operator[](lin_id))
Return a reference to the selected element.
Transform the boost::fusion::vector into memory specification (memory_traits)
__host__ static __device__ auto get_c(const data_type &data_, const g1_type &g1, const key_type &v1) -> decltype(boost::fusion::at_c< p >(data_.mem_r.operator[](g1.LinId(v1)))) &
Return a reference to the selected element.
v_transform_impl< H, first, last, exit_::value >::type type
generate the boost::fusion::vector apply H on each term
memory_traits_lin_type< T, openfpm::vect_isel< T >::value==OPENFPM_NATIVE >::type type
for each element in the vector interleave memory_c
static void deviceToHost(data_type &data_, size_t start, size_t stop)
Synchronize the memory buffer in the device with the memory in the host.
int yes_is_inte
indicate that it change the memory layout from the original
Definition: memory_conf.hpp:89
static void hostToDevice(data_type &data_, size_t start, size_t stop)
Copy the memory from host to device.
__host__ static __device__ auto get_lin_c(const data_type &data_, const g1_type &g1, const size_t lin_id) -> decltype(boost::fusion::at_c< p >(data_.mem_r.operator[](lin_id))) &
Return a reference to the selected element.
__host__ static __device__ auto get_c(const data_type &data_, const g1_type &g1, const key_type &v1) -> decltype(boost::fusion::at_c< p >(data_).mem_r.operator[](g1.LinId(v1)))
Return a reference to the selected element.
Transform the boost::fusion::vector into memory specification (memory_traits)
Definition: memory_conf.hpp:83
This class convert a boost::mpl::fusion/vector to a boost::mpl::fusion/vector with memory_c<....
Definition: memory_conf.hpp:63
static void hostToDevice(data_type &data_, size_t start, size_t stop)
Synchronize the memory buffer in the device with the memory in the host.
inter_memc< typename T::type >::type type
for each element in the vector interleave memory_c
Definition: memory_conf.hpp:86
static void deviceToHost(data_type &data_, size_t start, size_t stop)
Synchronize the memory buffer in the device with the memory in the host.
is_multiple_buffer_each_prp
this class is a functor for "for_each" algorithm
__host__ static __device__ auto get_lin(data_type &data_, const g1_type &g1, size_t lin_id) -> decltype(boost::fusion::at_c< p >(data_).mem_r.operator[](lin_id))
Return a reference to the selected element.
__host__ static __device__ auto get(data_type &data_, const g1_type &g1, const key_type &v1) -> decltype(boost::fusion::at_c< p >(data_.mem_r.operator[](g1.LinId(v1)))) &
Return a reference to the selected element.
KeyT const ValueT ValueT OffsetIteratorT OffsetIteratorT int
[in] The number of segments that comprise the sorting data
This class convert a boost::mpl::fusion/vector to a boost::mpl::fusion/vector with memory_c interleav...
Definition: memory_conf.hpp:40
Void structure.
Definition: common.hpp:73
__host__ static __device__ auto get(data_type &data_, const g1_type &g1, const key_type &v1) -> decltype(boost::fusion::at_c< p >(data_).mem_r.operator[](g1.LinId(v1)))
Return a reference to the selected element.
this class is a functor for "for_each" algorithm