1 | /* |
2 | * memory_stride.hpp |
3 | * |
4 | * Created on: Aug 17, 2014 |
5 | * Author: Pietro Incardona |
6 | */ |
7 | |
8 | #ifndef MEMORY_ARRAY_HPP_ |
9 | #define MEMORY_ARRAY_HPP_ |
10 | |
11 | #include "memory/memory.hpp" |
12 | |
13 | #include "util/cuda_util.hpp" |
14 | |
15 | /*! |
16 | * |
17 | * \brief This class give a representation to a chunk or memory |
18 | * |
19 | * This class give a representation to a chunk of memory as an array of T objects |
20 | * |
21 | * \param T This is the object type |
22 | * |
23 | */ |
24 | |
25 | template<typename T> |
26 | class memory_array |
27 | { |
28 | |
29 | #if defined(__GNUG__) || defined(__clang__) |
30 | |
31 | //! Internal pointer |
32 | T __attribute__((aligned(16))) * ptr; |
33 | |
34 | #else |
35 | |
36 | //! Internal pointer |
37 | T * ptr; |
38 | |
39 | #endif |
40 | |
41 | //! number of elements |
42 | size_t sz; |
43 | |
44 | /*! \brief Get the i element |
45 | * |
46 | * \param i element |
47 | * |
48 | * \return the element i |
49 | * |
50 | */ |
51 | T get(mem_id i) |
52 | { |
53 | return ptr[i]; |
54 | } |
55 | |
56 | |
57 | public: |
58 | |
59 | //! type returned by this structure |
60 | typedef T value_type; |
61 | |
62 | /*! \brief Initialize the memory array |
63 | * |
64 | * \param ptr pointer |
65 | * \param sz number of elements in the array |
66 | * \param init indicate if you have to initialize the memory |
67 | * |
68 | * \return the element i |
69 | * |
70 | */ |
71 | void initialize(void * ptr, size_t sz, bool init) |
72 | { |
73 | this->ptr = static_cast<T *>(ptr); |
74 | |
75 | |
76 | // Initialize the constructors |
77 | |
78 | if (init == false) |
79 | new (ptr)T[sz]; |
80 | |
81 | this->sz = sz; |
82 | } |
83 | |
84 | //! return the size of the array |
85 | size_t size() |
86 | { |
87 | return sz; |
88 | } |
89 | |
90 | /*! \brief Set the internal pointer to the indicated chunk of memory |
91 | * |
92 | * \param ptr_ pointer |
93 | * |
94 | */ |
95 | void set_pointer(void * ptr_) |
96 | { |
97 | ptr = static_cast<T *>(ptr_); |
98 | } |
99 | |
100 | //! Return the pointer |
101 | __device__ __host__ void * get_pointer() const |
102 | { |
103 | return ptr; |
104 | } |
105 | |
106 | /*! \brief Set from another memory array |
107 | * |
108 | * \param the other memory_array |
109 | * |
110 | */ |
111 | template<typename Tbind> void bind_ref(const memory_array<Tbind> & ref) |
112 | { |
113 | ptr = static_cast<T *>(ref.get_pointer()); |
114 | } |
115 | |
116 | /*! \brief Access element an element of the array |
117 | * |
118 | * \param i element to access |
119 | * |
120 | * \return a teference to the object |
121 | * |
122 | */ |
123 | __host__ __device__ T & operator[](mem_id i) |
124 | { |
125 | return ptr[i]; |
126 | } |
127 | |
128 | /*! \brief Access element an element of the array |
129 | * |
130 | * \param i element to access |
131 | * |
132 | * \return a teference to the object |
133 | * |
134 | */ |
135 | __host__ __device__ const T & operator[](mem_id i) const |
136 | { |
137 | return ptr[i]; |
138 | } |
139 | |
140 | /*! \brief swap the two objects memory |
141 | * |
142 | * \param obj memory to swap with |
143 | * |
144 | */ |
145 | void swap(memory_array<T> & obj) |
146 | { |
147 | size_t sz_tmp = sz; |
148 | sz = obj.sz; |
149 | obj.sz = sz_tmp; |
150 | |
151 | T * ptr_tmp = ptr; |
152 | ptr = obj.ptr; |
153 | obj.ptr = ptr_tmp; |
154 | } |
155 | |
156 | /*! \brief Deinitialize the memory |
157 | * |
158 | * |
159 | * |
160 | */ |
161 | void deinit() |
162 | { |
163 | // Call the destructor of every objects |
164 | for (size_t i = 0 ; i < sz ; i++) |
165 | { |
166 | (&ptr[i])->~T(); |
167 | } |
168 | } |
169 | |
170 | //! Default constructor |
171 | memory_array() |
172 | :ptr(NULL),sz(0) |
173 | {}; |
174 | |
175 | /*! \brief Memory array constructor |
176 | * |
177 | * \param ptr Memory pointer |
178 | * \param sz size |
179 | * \param init specify if the pointer is initialized |
180 | * |
181 | */ |
182 | memory_array(void * ptr, size_t sz, bool init) |
183 | { |
184 | initialize(ptr,sz,init); |
185 | } |
186 | |
187 | /*! \brief Destructor |
188 | * |
189 | * |
190 | * |
191 | */ |
192 | ~memory_array() |
193 | { |
194 | }; |
195 | }; |
196 | |
197 | |
198 | #endif |
199 | |