1/*
2 * memory_c.hpp
3 *
4 * Created on: Aug 17, 2014
5 * Author: Pietro Incardona
6 */
7
8#include <boost/multi_array.hpp>
9#include <boost/fusion/mpl.hpp>
10#include <boost/fusion/include/mpl.hpp>
11#include <boost/mpl/vector.hpp>
12#include <array>
13#include <boost/mpl/pop_front.hpp>
14#include <boost/mpl/push_front.hpp>
15
16//#include "util/boost/boost_multi_array_openfpm.hpp"
17#include "util/multi_array_openfpm/multi_array_ref_openfpm.hpp"
18#include "util/ct_array.hpp"
19#include "memory_array.hpp"
20#include "memory/memory.hpp"
21
22#ifndef MEMORY_C_HPP_
23#define MEMORY_C_HPP_
24
25#define MEMORY_C_STANDARD 1
26#define MEMORY_C_REDUCED 2
27
28template<typename T, unsigned int impl = MEMORY_C_STANDARD, typename D = memory>
29class memory_c
30{
31
32};
33
34/*!
35 * \brief This class is a container for the memory interface like HeapMemory CudaMemory
36 *
37 * It store the object used to allocate memory and a representation of this memory as an array of objects T
38 *
39 * It is mainly used by memory_conf to create the correct memory layout
40 *
41 * \see memory_traits_inte memory_traits_lin
42 *
43 */
44template<typename T, typename D>
45class memory_c<T,MEMORY_C_STANDARD,D>
46{
47 public:
48
49 //! define T
50 typedef memory_c<T> type;
51
52 //! define a reference to T
53 typedef T& reference;
54
55 //! define T
56 typedef T vtype;
57
58 //! object that allocate memory like HeapMemory or CudaMemory
59 memory * mem;
60
61 //! object that represent the memory as an array of objects T
62 memory_array<T> mem_r;
63
64 /*! \brief This function set the object that allocate memory
65 *
66 * \param mem the memory object
67 *
68 */
69 void setMemory(memory & mem)
70 {
71 if (this->mem != NULL)
72 {
73 this->mem->decRef();
74
75 if (this->mem->ref() == 0 && &mem != this->mem)
76 delete(this->mem);
77 }
78// this->mem->decRef();
79 mem.incRef();
80 this->mem = &mem;
81 }
82
83 /*! \brief This function bind the memory_c to this memory_c as reference
84 *
85 * Bind ad reference it mean this this object does not create new memory but use the one from ref
86 * as a reference.
87 *
88 */
89 bool bind_ref(const memory_c<T,MEMORY_C_STANDARD,D> & ref)
90 {
91 mem = ref.mem;
92 mem->incRef();
93
94 //! we create the representation for the memory buffer
95 mem_r = ref.mem_r;
96
97 return true;
98 }
99
100 /*! \brief This function get the object that allocate memory
101 *
102 * \return memory object to allocate memory
103 *
104 */
105
106 memory& getMemory()
107 {
108 return *this->mem;
109 }
110
111 /*! \brief Switch the pointer to device pointer
112 *
113 */
114 void switchToDevicePtr()
115 {
116 mem_r.set_pointer(mem->getDevicePointer());
117 }
118
119 /*! \brief This function get the object that allocate memory
120 *
121 * \return memory object to allocate memory
122 *
123 */
124
125 const memory& getMemory() const
126 {
127 return *this->mem;
128 }
129
130 /*! \brief This function allocate memory
131 *
132 */
133 bool allocate(const size_t sz, bool skip_initialization = false)
134 {
135 memory * mem = this->mem;
136
137 //! We create a chunk of memory
138 mem->resize( sz*sizeof(T) );
139
140 //! we create the representation for this buffer
141 mem_r.initialize(mem->getPointer(),sz,mem->isInitialized() | skip_initialization);
142
143 return true;
144 }
145
146 //! constructor
147 memory_c():mem(NULL){}
148
149 //! destructor
150 ~memory_c()
151 {
152 // deinitialixe mem_r
153 mem_r.deinit();
154 if (mem != NULL)
155 {
156 mem->decRef();
157
158 if (mem->ref() == 0)
159 delete(mem);
160 }
161 }
162
163 /*! \brief swap the memory
164 *
165 * swap the memory between objects
166 *
167 */
168 void swap(memory_c & mem_obj)
169 {
170 // Save on temporal
171
172 void * mem_tmp = static_cast<void*>(mem);
173 mem = mem_obj.mem;
174 mem_obj.mem = static_cast<memory*>(mem_tmp);
175
176 mem_obj.mem_r.swap(mem_r);
177 }
178
179 /*! \brief swap the memory
180 *
181 * swap the memory between objects
182 *
183 */
184 template<typename Mem_type>
185 __host__ void swap_nomode(memory_c & mem_obj)
186 {
187 // It would be better a dynamic_cast, unfortunately nvcc
188 // does not accept it. It seems that for some reason nvcc want to
189 // produce device code out of this method. While it is true
190 // that this method is called inside a generic __device__ __host__
191 // function, tagging this method only __host__ does not stop
192 // nvcc from the intention to produce device code.
193 // The workaround is to use static_cast. Another workaround (to test)
194 // could be create a duplicate for_each function tagged only __host__ ,
195 // but I have to intention to duplicate code
196
197 Mem_type * mem_tmp = static_cast<Mem_type*>(mem);
198 mem_tmp->swap(*static_cast<Mem_type*>(mem_obj.mem));
199
200 mem_obj.mem_r.swap(mem_r);
201 }
202
203};
204
205/*! \brief This class is a trick to indicate the compiler a specific
206 * specialization pattern
207 *
208 * In particular it say that a multidimensional array has been found and
209 * need a special treatment, T is suppose to be a boost::mpl::vector of
210 * unsigned int indicating each dimension. T has to be a type of known size at
211 * compile time
212 *
213 */
214template<typename T>
215class multi_array
216{
217 typedef T type;
218};
219
220
221/*! \brief this class multiply all the elements in a boost::mpl::vector excluding the first element
222 *
223 * \param T expecting a boost::mpl::vector
224 *
225 */
226template<typename T, unsigned int N>
227struct mult
228{
229 enum { value = mult<T,N-1>::value * boost::mpl::at<T,boost::mpl::int_<N>>::type::value };
230};
231
232template <typename T>
233struct mult<T,1>
234{
235 enum { value = boost::mpl::at<T,boost::mpl::int_<1>>::type::value };
236};
237
238template<typename size_type, unsigned int dim>
239static inline std::array<size_type ,dim> zero_dims()
240{
241 std::array<size_type ,dim> dimensions;
242
243 // fill runtime, and the other dimensions
244 for (size_t i = 0 ; i < dim ; i++)
245 {dimensions[i] = 0;}
246
247 return dimensions;
248}
249
250template<unsigned int dim, typename size_type>
251struct array_ord
252{
253};
254
255template<typename size_type>
256struct array_ord<4,size_type>
257{
258 static constexpr size_type data[4] = {0,3,2,1};
259};
260
261template<typename size_type>
262struct array_ord<3,size_type>
263{
264 static constexpr size_type data[3] = {0,2,1};
265};
266
267template<typename size_type>
268struct array_ord<2,size_type>
269{
270 static constexpr size_type data[2] = {0,1};
271};
272
273template<unsigned int dim>
274struct array_asc
275{
276};
277
278template<>
279struct array_asc<4>
280{
281 static constexpr bool data[4] = {true,true,true,true};
282};
283
284template<>
285struct array_asc<3>
286{
287 static constexpr bool data[3] = {true,true,true};
288};
289
290template<>
291struct array_asc<2>
292{
293 static constexpr bool data[2] = {true,true};
294};
295
296
297/*! \brief Specialization of memory_c for multi_array
298 *
299 * Specialization of memory_c for multi_array
300 *
301 * It is mainly used by memory_conf to create the correct layout
302 *
303 * \see memory_traits_inte memory_traits_lin
304 *
305 * \tparam T is suppose to be a boost::mpl::vector specifing at position 0 the type and at
306 * position 1 to N the dimensions size of the multi_array
307 *
308 * \tparam D object that allocate memory
309 *
310 */
311template<typename T, typename D>
312class memory_c<multi_array<T>, MEMORY_C_STANDARD, D>
313{
314 /*! \brief In combination with generate_array is used to produce array at compile-time
315 *
316 * In combination with generate_array is used to produce at compile-time
317 * arrays like {true,true,.........true} used in boost::multi_array to
318 * define ascending order
319 *
320 */
321 template<size_t index,size_t N> struct ascending
322 {
323 enum { value = true };
324 };
325
326 //! Remove the first element
327 typedef typename boost::mpl::push_front<typename boost::mpl::pop_front<T>::type,boost::mpl::int_<-1>>::type Tv;
328
329 //! define boost::mpl::int_ without boost::mpl
330 template<int S> using int_ = boost::mpl::int_<S>;
331
332 //! define the template vector size it give a number at compile time
333 typedef typename boost::mpl::size<T> size_p;
334
335 //! Define "at" meta function without boost::mpl
336 template< typename S, unsigned int n> using at = boost::mpl::at<S,boost::mpl::int_<n> >;
337
338 typedef typename at<T,0>::type base;
339
340 //! define size_type
341 typedef typename openfpm::multi_array_ref_openfpm<base,size_p::value,Tv>::size_type size_type;
342
343 public:
344
345 /*! \brief This function set the object that allocate memory
346 *
347 * \param mem the memory object
348 *
349 */
350
351 void setMemory(memory & mem)
352 {
353 if (this->mem != NULL)
354 {
355 this->mem->decRef();
356
357 if (this->mem->ref() == 0 && &mem != this->mem)
358 delete(this->mem);
359 }
360 mem.incRef();
361 this->mem = &mem;
362 }
363
364 /*! \brief This function get the object that allocate memory
365 *
366 * \return memory object to allocate memory
367 *
368 */
369
370 memory& getMemory()
371 {
372 return *this->mem;
373 }
374
375 /*! \brief Switch the pointer to device pointer
376 *
377 */
378 void switchToDevicePtr()
379 {
380 mem_r.set_pointer(mem->getDevicePointer());
381 }
382
383
384 /*! \brief This function bind the memory_c to this memory_c as reference
385 *
386 * Bind ad reference it mean this this object does not create new memory but use the one from ref
387 * as a reference.
388 *
389 */
390 bool bind_ref(const memory_c<multi_array<T>, MEMORY_C_STANDARD, D> & ref)
391 {
392 mem = ref.mem;
393 mem->incRef();
394
395 //! we create the representation for the memory buffer
396 mem_r = ref.mem_r;
397
398 return true;
399 }
400
401 /*! \brief This function allocate memory and associate the representation to mem_r
402 *
403 * This function allocate memory and associate the representation of that chunk of
404 * memory to mem_r
405 *
406 */
407 bool allocate(const size_t sz, bool skip_initialization = false)
408 {
409 memory * mem = this->mem;
410
411 //! We create a chunk of memory
412 mem->resize( sz*mult<T,size_p::value-1>::value*sizeof(base) );
413
414 openfpm::multi_array_ref_openfpm<base,size_p::value,Tv> tmp(static_cast<base *>(mem->getPointer()),
415 sz,
416 openfpm::general_storage_order<size_p::value>(openfpm::ofp_storage_order()));
417
418 //! we create the representation for the memory buffer
419 mem_r.swap(tmp);
420
421 return true;
422 }
423
424 //! define the type of the multi_array vector minus one of the original size
425 //! basically we remove the index 0 of the multi_array
426 typedef boost::multi_array<base,size_p::value> type;
427
428 //! Reference to an object to allocate memory
429 D * mem;
430
431 //! object that represent the memory as a multi-dimensional array of objects T
432 openfpm::multi_array_ref_openfpm<base,boost::mpl::size<T>::value,Tv> mem_r;
433
434 //! constructor
435 memory_c()
436 :mem(NULL),
437 mem_r(static_cast<base *>(NULL),0,openfpm::ofp_storage_order())
438 {}
439
440 //! destructor
441 ~memory_c()
442 {
443 if (mem != NULL)
444 {
445 mem->decRef();
446 if (mem->ref() == 0)
447 delete(mem);
448 }
449 }
450
451 //! set the device memory interface, the object that allocate memory
452 void set_mem(memory & mem)
453 {
454 this->mem = &mem;
455 }
456
457 /*! \brief swap the memory
458 *
459 * swap the memory between objects
460 *
461 */
462 void swap(memory_c & mem_obj)
463 {
464 // Save on temporal
465
466 void * mem_tmp = static_cast<void*>(mem);
467 mem = mem_obj.mem;
468 mem_obj.mem = static_cast<D*>(mem_tmp);
469
470 mem_r.swap(mem_obj.mem_r);
471 }
472
473
474 /*! \brief swap the memory
475 *
476 * While the previous one swap the mem the swap_nomode call the swap member of the mem object
477 *
478 * \note calling the swap member require knowledge of the object type, we cannot work on abtsract objects
479 *
480 * swap the memory between objects
481 *
482 */
483 template<typename Mem_type>
484 __host__ void swap_nomode(memory_c & mem_obj)
485 {
486 // It would be better a dynamic_cast, unfortunately nvcc
487 // does not accept it. It seems that for some reason nvcc want to
488 // produce device code out of this method. While it is true
489 // that this function is called inside a generic __device__ __host__
490 // function, tagging this method only __host__ does not stop
491 // nvcc from the intention to produce device code.
492 // The workaround is to use static_cast. Another workaround (to test)
493 // could be create a duplicate for_each function tagged only __host__ ,
494 // but I have to intention to duplicate code
495
496 Mem_type * mem_tmp = static_cast<Mem_type*>(mem);
497 mem_tmp->swap(*static_cast<Mem_type*>(mem_obj.mem));
498
499 mem_obj.mem_r.swap(mem_r);
500 }
501};
502
503//! Partial specialization for scalar N=0
504template<typename T>
505struct array_to_vmpl
506{
507};
508
509
510//! Partial specialization for N=1
511template<typename T,size_t N1>
512struct array_to_vmpl<T[N1]>
513{
514 //! the internal array primitive information represented into a boost mpl vector
515 typedef boost::mpl::vector<T,boost::mpl::int_<N1>> prim_vmpl;
516};
517
518
519//! Partial specialization for N=2
520template<typename T,size_t N1,size_t N2>
521struct array_to_vmpl<T[N1][N2]>
522{
523 //! the internal array primitive information represented into a boost mpl vector
524 typedef boost::mpl::vector<T,boost::mpl::int_<N1>,
525 boost::mpl::int_<N2>> prim_vmpl;
526};
527
528/*! \brief OpenFPM use memory_c<multi_array<T> ..... > to implement the structure of array layout
529 *
530 * This mean that the object returned by mem_r are complex objects that represent the memory view, these view has
531 * the purpose to hook at compile time the operator[] to give the feeling of using an array
532 *
533 * This view depend from the template parameter Tv in the member mem_r, that as you can see is difficult to reconstruct
534 * In some case deduction does not work because too complex. So we have to compute this type. this function does this
535 * given a type like float[3], it produce the Tv parameter
536 *
537 *
538 */
539template<typename T>
540struct to_memory_multi_array_ref_view
541{
542 // first we convert the type into a boost vector containing the primitive, followed by array dimension
543 typedef typename array_to_vmpl<T>::prim_vmpl prim_vmpl;
544
545 // Than we operate at compile-time the same operation memory_c<multi_array does
546 //! Remove the first element (this is the Tv parameter of )
547 typedef typename boost::mpl::push_front<typename boost::mpl::pop_front<prim_vmpl>::type,boost::mpl::int_<-1>>::type vmpl;
548
549
550};
551
552#endif
553
554