Block-Structured AMR Software Framework
Loading...
Searching...
No Matches
AMReX_TableData.H
Go to the documentation of this file.
1#ifndef AMREX_TABLE_DATA_H_
2#define AMREX_TABLE_DATA_H_
3#include <AMReX_Config.H>
4
5#include <AMReX.H>
6#include <AMReX_Array.H>
8#include <AMReX_GpuDevice.H>
9#include <AMReX_GpuPrint.H>
10
11#include <cstring>
12#include <iostream>
13#include <sstream>
14#include <type_traits>
15
16namespace amrex {
17
18template <typename T, typename IDX = int>
19struct Table1D
20{
21 T* AMREX_RESTRICT p = nullptr;
22 IDX begin = 1;
23 IDX end = 0;
24
25 constexpr Table1D () noexcept = default;
26
27 template <class U=T>
28 requires (std::is_const_v<U>)
30 constexpr Table1D (Table1D<std::remove_const_t<T>> const& rhs) noexcept
31 : p(rhs.p),
32 begin(rhs.begin),
33 end(rhs.end)
34 {}
35
37 constexpr Table1D (T* a_p, IDX a_begin, IDX a_end) noexcept
38 : p(a_p),
39 begin(a_begin),
40 end(a_end)
41 {}
42
44 explicit operator bool () const noexcept { return p != nullptr; }
45
47 T& operator() (IDX i) const noexcept
48 requires (!std::is_void_v<T>)
49 {
50#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
51 index_assert(i);
52#endif
53 return p[i-begin];
54 }
55
56#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
58 void index_assert (IDX i) const
59 {
60 if (i < begin || i >= end) {
61 if constexpr (std::is_same_v<IDX,int>) {
63 AMREX_DEVICE_PRINTF(" (%d) is out of bound (%d:%d)\n",
64 i, begin, end-1);
66 ))
67 } else if constexpr (std::is_same_v<IDX,long>) {
69 AMREX_DEVICE_PRINTF(" (%ld) is out of bound (%ld:%ld)\n",
70 i, begin, end-1);
72 ))
73 } else if constexpr (std::is_same_v<IDX,long long>) {
75 AMREX_DEVICE_PRINTF(" (%lld) is out of bound (%lld:%lld)\n",
76 i, begin, end-1);
78 ))
79 } else {
80 AMREX_IF_ON_DEVICE(( amrex::Abort(" Out of bound\n"); ))
81 }
83 std::stringstream ss;
84 ss << " (" << i << ") is out of bound ("
85 << begin << ":" << end-1 << ")";
86 amrex::Abort(ss.str());
87 ))
88 }
89 }
90#endif
91};
92
93template <typename T, Order ORDER = Order::F>
94struct Table2D
95{
96 T* AMREX_RESTRICT p = nullptr;
100
101 constexpr Table2D () noexcept = default;
102
103 template <class U=T>
104 requires (std::is_const_v<U>)
106 constexpr Table2D (Table2D<std::remove_const_t<T>, ORDER> const& rhs) noexcept
107 : p(rhs.p),
108 stride1(rhs.stride1),
109 begin(rhs.begin),
110 end(rhs.end)
111 {}
112
114 constexpr Table2D (T* a_p,
115 GpuArray<int,2> const& a_begin,
116 GpuArray<int,2> const& a_end) noexcept
117 : p(a_p),
118 stride1(len0(a_begin,a_end)),
119 begin(a_begin),
120 end(a_end)
121 {}
122
124 explicit operator bool () const noexcept { return p != nullptr; }
125
127 T& operator() (int i, int j) const noexcept
128 requires (!std::is_void_v<T>)
129 {
130#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
131 index_assert(i,j);
132#endif
133 if constexpr (ORDER == Order::F) {
134 return p[(i-begin[0])+(j-begin[1])*stride1];
135 } else {
136 return p[(i-begin[0])*stride1+(j-begin[1])];
137 }
138 }
139
140#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
142 void index_assert (int i, int j) const
143 {
144 if (i < begin[0] || i >= end[0] ||
145 j < begin[1] || j >= end[1]) {
147 AMREX_DEVICE_PRINTF(" (%d,%d) is out of bound (%d:%d,%d:%d)\n",
148 i, j, begin[0], end[0]-1, begin[1], end[1]-1);
149 amrex::Abort();
150 ))
152 std::stringstream ss;
153 ss << " (" << i << "," << j << ") is out of bound ("
154 << begin[0] << ":" << end[0]-1
155 << "," << begin[1] << ":" << end[1]-1 << ")";
156 amrex::Abort(ss.str());
157 ))
158 }
159 }
160#endif
161
162private:
163
164 static constexpr int len0 (GpuArray<int,2> const& a_begin,
165 GpuArray<int,2> const& a_end) noexcept
166 {
167 if constexpr (ORDER == Order::F) {
168 return a_end[0] - a_begin[0];
169 } else {
170 return a_end[1] - a_begin[1];
171 }
172 }
173};
174
175template <typename T, Order ORDER = Order::F>
177{
178 T* AMREX_RESTRICT p = nullptr;
179 Long stride1 = 0;
180 Long stride2 = 0;
183
184 constexpr Table3D () noexcept = default;
185
186 template <class U=T>
187 requires (std::is_const_v<U>)
189 constexpr Table3D (Table3D<std::remove_const_t<T>,ORDER> const& rhs) noexcept
190 : p(rhs.p),
191 stride1(rhs.stride1),
192 stride2(rhs.stride2),
193 begin(rhs.begin),
194 end(rhs.end)
195 {}
196
198 constexpr Table3D (T* a_p,
199 GpuArray<int,3> const& a_begin,
200 GpuArray<int,3> const& a_end) noexcept
201 : p(a_p),
202 stride1( len0(a_begin,a_end)),
203 stride2(stride1*len1(a_begin,a_end)),
204 begin(a_begin),
205 end(a_end)
206 {}
207
209 explicit operator bool () const noexcept { return p != nullptr; }
210
212 T& operator() (int i, int j, int k) const noexcept
213 requires (!std::is_void_v<T>)
214 {
215#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
216 index_assert(i,j,k);
217#endif
218 if constexpr (ORDER == Order::F) {
219 return p[(i-begin[0])+(j-begin[1])*stride1+(k-begin[2])*stride2];
220 } else {
221 return p[(i-begin[0])*stride2+(j-begin[1])*stride1+(k-begin[2])];
222 }
223 }
224
225#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
227 void index_assert (int i, int j, int k) const
228 {
229 if (i < begin[0] || i >= end[0] ||
230 j < begin[1] || j >= end[1] ||
231 k < begin[2] || k >= end[2]) {
233 AMREX_DEVICE_PRINTF(" (%d,%d,%d) is out of bound (%d:%d,%d:%d,%d:%d)\n",
234 i, j, k, begin[0], end[0]-1, begin[1], end[1]-1,
235 begin[2], end[2]-1);
236 amrex::Abort();
237 ))
239 std::stringstream ss;
240 ss << " (" << i << "," << j << "," << k << ") is out of bound ("
241 << begin[0] << ":" << end[0]-1
242 << "," << begin[1] << ":" << end[1]-1
243 << "," << begin[2] << ":" << end[2]-1 << ")";
244 amrex::Abort(ss.str());
245 ))
246 }
247 }
248#endif
249
250private:
251
252 static constexpr int len0 (GpuArray<int,3> const& a_begin,
253 GpuArray<int,3> const& a_end) noexcept
254 {
255 if constexpr (ORDER == Order::F) {
256 return a_end[0] - a_begin[0];
257 } else {
258 return a_end[2] - a_begin[2];
259 }
260 }
261
262 static constexpr int len1 (GpuArray<int,3> const& a_begin,
263 GpuArray<int,3> const& a_end) noexcept
264 {
265 return a_end[1] - a_begin[1];
266 }
267};
268
269template <typename T, Order ORDER = Order::F>
271{
272 T* AMREX_RESTRICT p = nullptr;
273 Long stride1 = 0;
274 Long stride2 = 0;
275 Long stride3 = 0;
278
279 constexpr Table4D () noexcept = default;
280
281 template <class U=T>
282 requires (std::is_const_v<U>)
284 constexpr Table4D (Table4D<std::remove_const_t<T>,ORDER> const& rhs) noexcept
285 : p(rhs.p),
286 stride1(rhs.stride1),
287 stride2(rhs.stride2),
288 stride3(rhs.stride3),
289 begin(rhs.begin),
290 end(rhs.end)
291 {}
292
294 constexpr Table4D (T* a_p,
295 GpuArray<int,4> const& a_begin,
296 GpuArray<int,4> const& a_end) noexcept
297 : p(a_p),
298 stride1( len0(a_begin,a_end)),
299 stride2(stride1*len1(a_begin,a_end)),
300 stride3(stride2*len2(a_begin,a_end)),
301 begin(a_begin),
302 end(a_end)
303 {}
304
306 explicit operator bool () const noexcept { return p != nullptr; }
307
309 T& operator() (int i, int j, int k, int n) const noexcept
310 requires (!std::is_void_v<T>)
311 {
312#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
313 index_assert(i,j,k,n);
314#endif
315 if constexpr (ORDER == Order::F) {
316 return p[(i-begin[0])+(j-begin[1])*stride1+(k-begin[2])*stride2+(n-begin[3])*stride3];
317 } else {
318 return p[(i-begin[0])*stride3+(j-begin[1])*stride2+(k-begin[2])*stride1+(n-begin[3])];
319 }
320 }
321
322#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
324 void index_assert (int i, int j, int k, int n) const
325 {
326 if (i < begin[0] || i >= end[0] ||
327 j < begin[1] || j >= end[1] ||
328 k < begin[2] || k >= end[2] ||
329 n < begin[3] || n >= end[3]) {
331 AMREX_DEVICE_PRINTF(" (%d,%d,%d,%d) is out of bound (%d:%d,%d:%d,%d:%d,%d:%d)\n",
332 i, j, k, n, begin[0], end[0]-1, begin[1], end[1]-1,
333 begin[2], end[2]-1, begin[3], end[3]-1);
334 amrex::Abort();
335 ))
337 std::stringstream ss;
338 ss << " (" << i << "," << j << "," << k << "," << n << ") is out of bound ("
339 << begin[0] << ":" << end[0]-1
340 << "," << begin[1] << ":" << end[1]-1
341 << "," << begin[2] << ":" << end[2]-1
342 << "," << begin[3] << ":" << end[3]-1 << ")";
343 amrex::Abort(ss.str());
344 ))
345 }
346 }
347#endif
348
349private:
350
351 static constexpr int len0 (GpuArray<int,4> const& a_begin,
352 GpuArray<int,4> const& a_end) noexcept
353 {
354 if constexpr (ORDER == Order::F) {
355 return a_end[0] - a_begin[0];
356 } else {
357 return a_end[3] - a_begin[3];
358 }
359 }
360
361 static constexpr int len1 (GpuArray<int,4> const& a_begin,
362 GpuArray<int,4> const& a_end) noexcept
363 {
364 if constexpr (ORDER == Order::F) {
365 return a_end[1] - a_begin[1];
366 } else {
367 return a_end[2] - a_begin[2];
368 }
369 }
370
371 static constexpr int len2 (GpuArray<int,4> const& a_begin,
372 GpuArray<int,4> const& a_end) noexcept
373 {
374 if constexpr (ORDER == Order::F) {
375 return a_end[2] - a_begin[2];
376 } else {
377 return a_end[1] - a_begin[1];
378 }
379 }
380};
381
420template <typename T, int N, Order ORDER = Order::F>
422 : public DataAllocator
423{
424public:
425
426 template <class U, int M, Order O> friend class TableData;
427 using value_type = T;
428 using table_type = std::conditional_t<N==1, Table1D<T>,
429 std::conditional_t<N==2, Table2D<T, ORDER>,
430 std::conditional_t<N==3, Table3D<T, ORDER>,
431 Table4D<T, ORDER> > > >;
432 using const_table_type = std::conditional_t<N==1, Table1D<T const>,
433 std::conditional_t<N==2, Table2D<T const, ORDER>,
434 std::conditional_t<N==3, Table3D<T const, ORDER>,
436
437 TableData () noexcept = default;
438
439 explicit TableData (Arena* ar) noexcept;
440
441 TableData (Array<int,N> const& lo, Array<int,N> const& hi, Arena* ar = nullptr);
442
443 TableData (TableData<T,N,ORDER> const&) = delete;
444 TableData<T,N,ORDER>& operator= (TableData<T,N,ORDER> const&) = delete;
445
446 TableData (TableData<T,N,ORDER>&& rhs) noexcept;
447 TableData<T,N,ORDER>& operator= (TableData<T,N,ORDER> && rhs) noexcept;
448
449 ~TableData () noexcept;
450
451 [[nodiscard]] constexpr int dim () const noexcept { return N; }
452
453 void resize (Array<int,N> const& lo, Array<int,N> const& hi, Arena* ar = nullptr);
454
455 [[nodiscard]] Long size () const noexcept;
456
457 Array<int,N> const& lo () const noexcept { return m_lo; }
458
459 Array<int,N> const& hi () const noexcept { return m_hi; }
460
461 void clear () noexcept;
462
463 void copy (TableData<T,N,ORDER> const& rhs) noexcept;
464
465 table_type table () noexcept;
466 const_table_type table () const noexcept;
467 const_table_type const_table () const noexcept;
468
469private:
470
471 void define ();
472
473 T* m_dptr = nullptr;
474 Array<int,N> m_lo;
475 Array<int,N> m_hi;
476 Long m_truesize = 0L;
477 bool m_ptr_owner = false;
478};
479
480template <typename T, int N, Order ORDER>
481TableData<T,N,ORDER>::TableData (Array<int,N> const& lo, Array<int,N> const& hi, Arena* ar)
482 : DataAllocator{ar}, m_lo(lo), m_hi(hi)
483{
484 define();
485}
486
487
488template <typename T, int N, Order ORDER>
490 : DataAllocator{rhs.arena()},
491 m_dptr(rhs.m_dptr),
492 m_lo(rhs.m_lo),
493 m_hi(rhs.m_hi),
494 m_truesize(rhs.m_truesize),
495 m_ptr_owner(rhs.m_ptr_owner)
496{
497 rhs.m_dptr = nullptr;
498 rhs.m_ptr_owner = false;
499}
500
501template <typename T, int N, Order ORDER>
502TableData<T,N,ORDER>&
504{
505 if (this != &rhs) {
506 clear();
507 m_arena = rhs.m_arena;
508 m_dptr = rhs.m_dptr;
509 m_lo = rhs.m_lo;
510 m_hi = rhs.m_hi;
511 m_truesize = rhs.m_truesize;
512 m_ptr_owner = rhs.m_ptr_owner;
513 rhs.m_dptr = nullptr;
514 rhs.m_ptr_owner = false;
515 }
516 return *this;
517}
518
519template <typename T, int N, Order ORDER>
521{
522 static_assert(std::is_trivially_copyable<T>() &&
523 std::is_trivially_destructible<T>(),
524 "TableData<T,N,ORDER>: T must be trivially copyable and trivially destructible");
525 static_assert(N>=1 && N <=4, "TableData<T,N,ORDER>: N must be in the range of [1,4]");
526 clear();
527}
528
529template <typename T, int N, Order ORDER>
530void
532{
533 m_lo = lo;
534 m_hi = hi;
535
536 if (ar == nullptr) {
537 ar = m_arena;
538 }
539
540 if (arena() != DataAllocator(ar).arena()) {
541 clear();
542 m_arena = ar;
543 define();
544 } else if (m_dptr == nullptr || !m_ptr_owner) {
545 m_dptr = nullptr;
546 define();
547 } else if (size() > m_truesize) {
548 clear();
549 define();
550 }
551}
552
553template <typename T, int N, Order ORDER>
554Long
556{
557 Long r = 1;
558 for (int i = 0; i < N; ++i) {
559 r *= m_hi[i] - m_lo[i] + 1;
560 }
561 return r;
562}
563
564template <typename T, int N, Order ORDER>
565void
567{
568 if (m_dptr) {
569 if (m_ptr_owner) {
570 this->free(m_dptr);
571 }
572 m_dptr = nullptr;
573 m_truesize = 0;
574 }
575}
576
577template <typename T, int N, Order ORDER>
578void
580{
581 m_truesize = size();
582 AMREX_ASSERT(m_truesize >= 0);
583 if (m_truesize == 0) {
584 return;
585 } else {
586 m_ptr_owner = true;
587 m_dptr = static_cast<T*>(this->alloc(m_truesize*sizeof(T)));
588 }
589}
590
592namespace detail {
593 template <typename T, Order>
594 Table1D<T> make_table (T* p, Array<int,1> const& lo, Array<int,1> const& hi) {
595 return Table1D<T>(p, lo[0], hi[0]+1);
596 }
597 template <typename T, Order ORDER>
598 Table2D<T,ORDER> make_table (T* p, Array<int,2> const& lo, Array<int,2> const& hi) {
599 return Table2D<T,ORDER>(p, {lo[0],lo[1]}, {hi[0]+1,hi[1]+1});
600 }
601 template <typename T, Order ORDER>
602 Table3D<T,ORDER> make_table (T* p, Array<int,3> const& lo, Array<int,3> const& hi) {
603 return Table3D<T,ORDER>(p, {lo[0],lo[1],lo[2]}, {hi[0]+1,hi[1]+1,hi[2]+1});
604 }
605 template <typename T, Order ORDER>
606 Table4D<T,ORDER> make_table (T* p, Array<int,4> const& lo, Array<int,4> const& hi) {
607 return Table4D<T,ORDER>(p, {lo[0],lo[1],lo[2],lo[3]}, {hi[0]+1,hi[1]+1,hi[2]+1,hi[3]+1});
608 }
609}
611
612template <typename T, int N, Order ORDER>
613typename TableData<T,N,ORDER>::table_type
615{
616 return detail::make_table<T,ORDER>(m_dptr, m_lo, m_hi);
617}
618
619template <typename T, int N, Order ORDER>
622{
623 return detail::make_table<T const, ORDER>(m_dptr, m_lo, m_hi);
624}
625
626template <typename T, int N, Order ORDER>
629{
630 return detail::make_table<T const, ORDER>(m_dptr, m_lo, m_hi);
631}
632
633template <typename T, int N, Order ORDER>
634void
636{
637 std::size_t count = sizeof(T)*size();
638 if (count == 0) { return; }
639 AMREX_ASSERT(count == sizeof(T)*rhs.size());
640#ifdef AMREX_USE_GPU
641 bool this_on_device = arena()->isManaged() || arena()->isDevice();
642 bool rhs_on_device = rhs.arena()->isManaged() || rhs.arena()->isDevice();
643 if (this_on_device && rhs_on_device) {
644 Gpu::dtod_memcpy_async(m_dptr, rhs.m_dptr, count);
645 } else if (this_on_device && !rhs_on_device) {
646 Gpu::htod_memcpy_async(m_dptr, rhs.m_dptr, count);
647 } else if (!this_on_device && rhs_on_device) {
648 Gpu::dtoh_memcpy_async(m_dptr, rhs.m_dptr, count);
649 } else
650#endif
651 {
652 std::memcpy(m_dptr, rhs.m_dptr, count);
653 }
654}
655
656}
657
658#endif
#define AMREX_ASSERT(EX)
Definition AMReX_BLassert.H:38
#define AMREX_FORCE_INLINE
Definition AMReX_Extension.H:119
#define AMREX_RESTRICT
Definition AMReX_Extension.H:32
#define AMREX_DEVICE_PRINTF(...)
Definition AMReX_GpuPrint.H:21
#define AMREX_IF_ON_DEVICE(CODE)
Definition AMReX_GpuQualifiers.H:56
#define AMREX_IF_ON_HOST(CODE)
Definition AMReX_GpuQualifiers.H:58
#define AMREX_GPU_HOST_DEVICE
Definition AMReX_GpuQualifiers.H:20
A virtual base class for objects that manage their own dynamic memory allocation.
Definition AMReX_Arena.H:132
Multi-dimensional array class.
Definition AMReX_TableData.H:423
std::conditional_t< N==1, Table1D< T >, std::conditional_t< N==2, Table2D< T, ORDER >, std::conditional_t< N==3, Table3D< T, ORDER >, Table4D< T, ORDER > > > > table_type
Definition AMReX_TableData.H:431
TableData() noexcept=default
Array< int, N > const & hi() const noexcept
Definition AMReX_TableData.H:459
std::conditional_t< N==1, Table1D< T const >, std::conditional_t< N==2, Table2D< T const, ORDER >, std::conditional_t< N==3, Table3D< T const, ORDER >, Table4D< T const, ORDER > > > > const_table_type
Definition AMReX_TableData.H:435
T value_type
Definition AMReX_TableData.H:427
amrex_long Long
Definition AMReX_INT.H:30
std::array< T, N > Array
Definition AMReX_Array.H:26
Definition AMReX_Amr.cpp:50
Order
Definition AMReX_Order.H:7
__host__ __device__ Dim3 begin(BoxND< dim > const &box) noexcept
Definition AMReX_Box.H:2018
void Abort(const std::string &msg)
Print out message to cerr and exit via abort().
Definition AMReX.cpp:241
__host__ __device__ Dim3 end(BoxND< dim > const &box) noexcept
Definition AMReX_Box.H:2028
Definition AMReX_DataAllocator.H:9
Arena * arena() const noexcept
Definition AMReX_DataAllocator.H:24
Fixed-size array that can be used on GPU.
Definition AMReX_Array.H:43
Definition AMReX_TableData.H:20
IDX begin
Definition AMReX_TableData.H:22
T *__restrict__ p
Definition AMReX_TableData.H:21
__host__ __device__ T & operator()(IDX i) const noexcept
Definition AMReX_TableData.H:47
IDX end
Definition AMReX_TableData.H:23
constexpr Table1D() noexcept=default
__host__ __device__ constexpr Table1D(T *a_p, IDX a_begin, IDX a_end) noexcept
Definition AMReX_TableData.H:37
Definition AMReX_TableData.H:95
__host__ __device__ T & operator()(int i, int j) const noexcept
Definition AMReX_TableData.H:127
T *__restrict__ p
Definition AMReX_TableData.H:96
constexpr Table2D() noexcept=default
__host__ __device__ constexpr Table2D(T *a_p, GpuArray< int, 2 > const &a_begin, GpuArray< int, 2 > const &a_end) noexcept
Definition AMReX_TableData.H:114
GpuArray< int, 2 > end
Definition AMReX_TableData.H:99
GpuArray< int, 2 > begin
Definition AMReX_TableData.H:98
Long stride1
Definition AMReX_TableData.H:97
Definition AMReX_TableData.H:177
__host__ __device__ constexpr Table3D(T *a_p, GpuArray< int, 3 > const &a_begin, GpuArray< int, 3 > const &a_end) noexcept
Definition AMReX_TableData.H:198
constexpr Table3D() noexcept=default
Definition AMReX_TableData.H:271
constexpr Table4D() noexcept=default
__host__ __device__ constexpr Table4D(T *a_p, GpuArray< int, 4 > const &a_begin, GpuArray< int, 4 > const &a_end) noexcept
Definition AMReX_TableData.H:294