Block-Structured AMR Software Framework
Loading...
Searching...
No Matches
AMReX_Array4.H
Go to the documentation of this file.
1#ifndef AMREX_ARRAY4_H_
2#define AMREX_ARRAY4_H_
3
4#include <AMReX_Config.H>
5
11#include <AMReX.H>
12#include <AMReX_IntVect.H>
13#include <AMReX_GpuPrint.H>
14#include <AMReX_ConstexprFor.H>
15
16#include <iostream>
17#include <sstream>
18
20#define AMREX_ARRAY4_INDEX_ASSERT(i,j,k,n) \
21 if ((i)<begin.vect[0] || (i)>=end.vect[0] || \
22 (j)<begin.vect[1] || (j)>=end.vect[1] || \
23 (k)<begin.vect[2] || (k)>=end.vect[2] || \
24 (n)<0 || (n)>=end.vect[3]) \
25 { \
26 index_assert_print_error_message(i,j,k,n); \
27 }
29
30namespace amrex {
31
32 template<int dim>
33 class BoxND;
34
45 template <typename T>
46 struct CellData
47 {
49 T* AMREX_RESTRICT p = nullptr;
50 Long stride = 0;
51 int ncomp = 0;
53
64 constexpr CellData (T* a_p, Long a_stride, int a_ncomp)
65 : p(a_p), stride(a_stride), ncomp(a_ncomp)
66 {}
67
76 template <class U=T>
77 requires (std::is_const_v<U>)
79 constexpr CellData (CellData<std::remove_const_t<T>> const& rhs) noexcept
80 : p(rhs.p), stride(rhs.stride), ncomp(rhs.ncomp)
81 {}
82
89 explicit operator bool() const noexcept { return p != nullptr; }
90
92 [[nodiscard]] AMREX_GPU_HOST_DEVICE
93 int nComp() const noexcept { return ncomp; }
94
106 T& operator[] (int n) const
107 requires (!std::is_void_v<T>)
108
109 {
110#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
111 if (n < 0 || n >= ncomp) {
113 AMREX_DEVICE_PRINTF(" %d is out of bound (0:%d)", n, ncomp-1);
114 amrex::Abort();
115 ))
117 std::stringstream ss;
118 ss << " " << n << " is out of bound: (0:" << ncomp-1 << ")";
119 amrex::Abort(ss.str());
120 ))
121 }
122#endif
123 return p[n*stride];
124 }
125 };
126
128 namespace detail {
129
130 template <int N> struct Stride { Long a[N] = {}; };
131 template <> struct Stride<0> {};
132
138 template <typename T>
139 struct IsValidIndexType {
140 static constexpr bool value = std::is_integral_v<T> && IsNonNarrowingConversion_v<T, int>;
141 };
142
149 template <int N, bool last_dim_component, typename... idx>
150 struct ArrayNDIndexCheck_impl {
151 private:
152 static constexpr int num_indices = sizeof...(idx);
153
154 template <typename... Ts>
155 struct AllExceptLastEnum;
156
157 template <typename first, typename... Ts>
158 struct AllExceptLastEnum<first, Ts...> {
159 static constexpr bool value =
160 IsValidIndexType<std::decay_t<first>>::value
161 && AllExceptLastEnum<Ts...>::value;
162 };
163
164 // Check if the last index type is an enum or valid index type.
165 // An enum is only allowed if last_dim_component is true.
166 template <typename last>
167 struct AllExceptLastEnum<last> {
168 static constexpr bool value = IsValidIndexType<std::decay_t<last>>::value
169 || (std::is_enum_v<std::decay_t<last>>);
170 };
171
172 static constexpr bool index_with_maybe_enum = AllExceptLastEnum<idx...>::value;
173
174 static constexpr bool index_all_values = Conjunction<
175 IsValidIndexType<std::decay_t<idx>>...>::value;
176
177 public:
178 static constexpr bool value = ((num_indices == N) && index_all_values) // N indices are integer types
179 || ((num_indices == N - 1) && index_all_values && last_dim_component) // N-1 indices are integer types, last is component
180 || ((num_indices == N) && index_with_maybe_enum && last_dim_component); // N indices, last dim is component and can be enum
181 };
182
183 template <int N, bool last_dim_component, class... idx>
184 inline constexpr bool ArrayNDIndexCheck_impl_v = ArrayNDIndexCheck_impl<N, last_dim_component, idx...>::value;
185
186 template<std::size_t... idx>
187 constexpr auto make_oob_message_impl (std::index_sequence<idx...>) {
188 constexpr std::size_t N = sizeof...(idx);
189 constexpr char prefix[] = " (";
190 constexpr char middle[] = ") is out of bound (";
191 constexpr char suffix[] = ")\n";
192
193 constexpr std::size_t size =
194 sizeof(prefix) - 1 +
195 (2*N + (N-1)) +
196 sizeof(middle) - 1 +
197 (5*N + (N-1)) +
198 sizeof(suffix); // include null terminator
199
200 std::array<char, size> buf{};
201
202 std::size_t pos = 0;
203
204 // prefix
205 for (char c : prefix) {
206 if (c != '\0') {
207 buf[pos++] = c;
208 }
209 }
210 // %d,%d,...
211 ((buf[pos++] = '%', buf[pos++] = 'd', idx + 1 < N ? buf[pos++] = ',' : 0), ...);
212
213 // ") is out of bound ("
214 for (char c : middle) {
215 if (c != '\0') {
216 buf[pos++] = c;
217 }
218 }
219
220 // %d:%d,%d:%d,...
221 ((buf[pos++] = '%', buf[pos++] = 'd', buf[pos++] = ':', buf[pos++] = '%', buf[pos++] = 'd',
222 idx + 1 < N ? buf[pos++] = ',' : 0), ...);
223
224 // suffix
225 for (char c : suffix) {
226 buf[pos++] = c;
227 }
228 return buf;
229 }
230
231 template <std::size_t... idx, std::size_t... idx2x>
233 void device_print_impl2 (std::index_sequence<idx...>,
234 std::index_sequence<idx2x...>,
235 IntVectND<sizeof...(idx)> const& iv,
236 IntVectND<sizeof...(idx)> const& begin,
237 IntVectND<sizeof...(idx)> const& end)
238 {
239 constexpr auto msg = make_oob_message_impl(std::index_sequence<idx...>{});
240
241 AMREX_DEVICE_PRINTF(msg.data(),
242 iv.vect[idx]...,
243 ((idx2x % 2 == 0) ? begin.vect[idx2x / 2] : end.vect[idx2x / 2] - 1)...
244 );
245 }
246
247 template <int N>
248 requires (N >= 1)
250 void device_printf_impl (const IntVectND<N>& iv,
251 const IntVectND<N>& begin,
252 const IntVectND<N>& end)
253 {
254 device_print_impl2(
255 std::make_index_sequence<N>{},
256 std::make_index_sequence<N*2>{},
257 iv, begin, end
258 );
259 }
260 }
262
287 template<typename T, int N, bool last_dim_component = false>
288 struct ArrayND
289 {
290 static_assert(N >= 1, "ArrayND must have at least one dimension");
291 static_assert(N > 1 || !last_dim_component, "ArrayND with N=1 cannot have last_dim_component=true");
292
294 static constexpr bool IsLastDimComponent_v = last_dim_component;
296 static constexpr bool IsArray4_v = (N==4 && last_dim_component);
297
298 T* AMREX_RESTRICT p = nullptr;
300 AMREX_NO_UNIQUE_ADDRESS detail::Stride<N-1> stride{};
304
311 constexpr ArrayND () noexcept : p(nullptr) {}
312
321 template <class U=T>
322 requires (std::is_const_v<U>)
324 constexpr ArrayND (ArrayND<std::remove_const_t<T>, N, last_dim_component> const& rhs) noexcept
325 : p(rhs.p), stride(rhs.stride), begin(rhs.begin), end(rhs.end) {}
326
332 // TODO: Make BoxND functions constexpr to allow this constructor to be constexpr.
334 ArrayND (T* a_p, BoxND<N> const& box) noexcept
335 requires (!last_dim_component)
336 : ArrayND(a_p, box.smallEnd(), box.bigEnd() + 1)
337 {}
338
345 template <int M>
346 requires (((M+1==N) || (N == 4 && M == AMREX_SPACEDIM))
347 && last_dim_component)
348 // TODO: Make BoxND functions constexpr to allow this constructor to be constexpr.
350 ArrayND (T* a_p, BoxND<M> const& box, int ncomp) noexcept
351 : ArrayND(a_p, box.smallEnd(), box.bigEnd() + 1, ncomp)
352 {}
353
364 constexpr ArrayND (T* a_p, IntVectND<N> const& a_begin, IntVectND<N> const& a_end) noexcept
365 requires (!last_dim_component)
366 : p(a_p), begin(a_begin), end(a_end)
367 {
368 set_stride();
369 }
370
382 constexpr ArrayND (T* a_p, Dim3 const& a_begin, Dim3 const& a_end, int a_ncomp) noexcept
383 requires (IsArray4_v)
384 : p(a_p), begin(a_begin.x, a_begin.y, a_begin.z, 0), end(a_end.x, a_end.y, a_end.z, a_ncomp)
385 {
386 set_stride();
387 }
388
399 template <int M>
400 requires (((M+1 == N) || (N == 4 && M == AMREX_SPACEDIM)) && last_dim_component)
402 constexpr ArrayND (T* a_p, IntVectND<M> const& a_begin, IntVectND<M> const& a_end, int ncomp) noexcept
403 : p(a_p)
404 {
405 constexpr_for<0, M>([&](int d) {
406 begin.vect[d] = a_begin.vect[d];
407 end.vect[d] = a_end.vect[d];
408 });
409 constexpr_for<M, N>([&](int d) {
410 begin.vect[d] = 0;
411 end.vect[d] = 1;
412 });
413 end.vect[N-1] = ncomp;
414
415 set_stride();
416 }
417
426 template <class U>
427 requires (std::is_same_v<std::remove_const_t<T>, std::remove_const_t<U>>
428 && (N >= 2) && last_dim_component)
430 constexpr ArrayND (ArrayND<U, N, last_dim_component> const& rhs, int start_comp) noexcept
431 : p((T*)(rhs.p + start_comp*rhs.stride.a[N-2])),
432 stride(rhs.stride),
433 begin(rhs.begin),
434 end(rhs.end)
435 {
436 begin.vect[N-1] = 0;
437 end.vect[N-1] = rhs.end.vect[N-1] - start_comp;
438 }
439
449 template <class U>
450 requires (std::is_same_v<std::remove_const_t<T>, std::remove_const_t<U>>
451 && (N >= 2) && last_dim_component)
453 constexpr ArrayND (ArrayND<U, N, last_dim_component> const& rhs, int start_comp, int num_comp) noexcept
454 : p((T*)(rhs.p + start_comp*rhs.stride.a[N-2])),
455 stride(rhs.stride),
456 begin(rhs.begin),
457 end(rhs.end)
458 {
459 begin.vect[N-1] = 0;
460 end.vect[N-1] = num_comp;
461 }
462
468 constexpr explicit operator bool() const noexcept { return p != nullptr; }
469
475 [[nodiscard]] AMREX_GPU_HOST_DEVICE
476 constexpr bool ok () const noexcept { return p != nullptr && end.allGT(begin); }
477
494 template <typename... idx>
495 requires (!std::is_void_v<T> && !IsArray4_v
496 && detail::ArrayNDIndexCheck_impl_v<N, last_dim_component, idx...>)
498 T& operator() (idx... i) const noexcept {
499 constexpr auto nidx = sizeof...(i);
500#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
501 index_assert(IntVectND<nidx>{i...});
502#endif
503 return p[get_offset(IntVectND<nidx>{i...})];
504 }
505
518 template <int M>
519 requires (!std::is_void_v<T> &&
520 ((M == N) || (!IsArray4_v && last_dim_component && (M + 1 == N))))
522 T& operator() (IntVectND<M> const& iv) const noexcept {
523#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
524 index_assert(iv);
525#endif
526 return p[get_offset(iv)];
527 }
528
540 template <int M>
541 requires (!std::is_void_v<T> && last_dim_component && !IsArray4_v
542 && (M + 1 == N))
544 T& operator() (IntVectND<M> const& iv, int n) const noexcept {
545#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
546 index_assert(iv, n);
547#endif
548 return p[get_offset(iv, n)];
549 }
550
557 template <typename... idx>
558 requires (!std::is_void_v<T> && !IsArray4_v
559 && detail::ArrayNDIndexCheck_impl_v<N, last_dim_component, idx...>)
561 T* ptr (idx... i) const noexcept {
562 constexpr auto nidx = sizeof...(i);
563#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
564 index_assert(IntVectND<nidx>{i...});
565#endif
566 return p + get_offset(IntVectND<nidx>{i...});
567 }
568
581 template <int M>
582 requires ((M == N) || (!IsArray4_v && last_dim_component && (M + 1 == N)))
584 T* ptr (IntVectND<M> const& iv) const noexcept {
585#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
586 index_assert(iv);
587#endif
588 return p + get_offset(iv);
589 }
590
602 template <int M>
603 requires (last_dim_component && !IsArray4_v && (M + 1 == N))
605 T* ptr (IntVectND<M> const& iv, int n) const noexcept {
606#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
607 index_assert(iv, n);
608#endif
609 return p + get_offset(iv, n);
610 }
611
617 constexpr T* dataPtr () const noexcept {
618 return this->p;
619 }
620
626 constexpr int nComp () const noexcept {
627 if constexpr (last_dim_component) {
628 return end.vect[N-1];
629 } else {
630 return 1;
631 }
632 }
633
643 constexpr std::size_t size () const noexcept {
644 if (ok()) {
645 std::size_t s = 1;
646 constexpr_for<0, N>([&](int d) {
647 s *= (end.vect[d] - begin.vect[d]);
648 });
649 return s;
650 } else {
651 return 0;
652 }
653 }
654
665 template <int d>
666 requires ((d < N) && (d >= 0))
668 constexpr Long get_stride () const noexcept {
669 if constexpr (N > 1 && d > 0) {
670 return stride.a[d-1];
671 } else {
672 return 1;
673 }
674 }
675
687 template <typename... idx>
688 requires (!IsArray4_v &&
689 detail::ArrayNDIndexCheck_impl_v<N, last_dim_component, idx...>)
691 constexpr bool contains (idx... i) const noexcept {
692 constexpr auto nidx = sizeof...(i);
693 return this->contains(IntVectND<nidx>{i...});
694 }
695
712 template <int M>
713 requires ((M == N) || (!IsArray4_v && last_dim_component && (M + 1 == N)))
715 constexpr bool contains (IntVectND<M> const& iv) const noexcept {
716 bool inside = true;
717 constexpr_for<0, M>([&](int d) {
718 inside = inside && (iv.vect[d] >= begin.vect[d]) && (iv.vect[d] < end.vect[d]);
719 });
720 return inside;
721 }
722
739 template <int M>
740 requires (last_dim_component && !IsArray4_v && (M + 1 == N))
742 constexpr bool contains (IntVectND<M> const& iv, int n) const noexcept {
743 bool inside = true;
744 constexpr_for<0, M>([&](int d) {
745 inside = inside && (iv.vect[d] >= begin.vect[d]) && (iv.vect[d] < end.vect[d]);
746 });
747 inside = inside && (n >= 0) && (n < end.vect[N-1]);
748 return inside;
749 }
750
768 CellData<T> cellData (int i, int j, int k) const noexcept
769 requires (IsArray4_v)
770
771 {
772 int ncomp = end.vect[N-1];
773 return CellData<T>(this->ptr(i,j,k), stride.a[N-2], ncomp);
774 }
775
783 template <int M>
784 requires ((M == N) || (last_dim_component && (M + 1 == N || M == AMREX_SPACEDIM)))
786 constexpr Long get_offset (IntVectND<M> const& iv) const noexcept
787 {
788 Long offset = iv.vect[0] - begin.vect[0];
789 // If M == N and we have a component at the end, we only loop up to N-1 for spatial.
790 // Otherwise, we loop up to M.
791 constexpr int idx = (last_dim_component && M == N) ? N - 1 : M;
792 if constexpr (N > 1) {
793 constexpr_for<1, idx>([&](int d) {
794 offset += (iv.vect[d] - begin.vect[d]) * stride.a[d-1];
795 });
796 }
797
798 // Handle the component offset if the input is the full N dimensions
799 // and the last dimension is a component.
800 if constexpr (last_dim_component && M == N) {
801 offset += iv.vect[N-1] * stride.a[N-2];
802 }
803 return offset;
804 }
805
815 template <int M>
816 requires (last_dim_component
817 && ((M + 1 == N) || (N == 4 && M == AMREX_SPACEDIM)))
819 constexpr Long get_offset (IntVectND<M> const& iv, int n) const noexcept
820 {
821 Long offset = iv.vect[0] - begin.vect[0];
822 constexpr_for<1, M>([&](int d) {
823 offset += (iv.vect[d] - begin.vect[d]) * stride.a[d-1];
824 });
825 offset += n * stride.a[N-2];
826 return offset;
827 }
828
829#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
830#if defined(AMREX_USE_HIP)
832#else
834#endif
835 void index_assert (IntVectND<N> const& iv) const
836 {
837 bool out_of_bounds = false;
838 for (int d = 0; d < N; ++d) {
839 if (iv.vect[d] < begin.vect[d] || iv.vect[d] >= end.vect[d]) {
840 out_of_bounds = true;
841 }
842 }
843 if (out_of_bounds) {
845 detail::device_printf_impl(iv, begin, end);
846 amrex::Abort();
847 ))
849 std::stringstream ss;
850 ss << " (";
851 for (int d = 0; d < N; ++d) {
852 ss << iv.vect[d];
853 if (d + 1 < N) ss << ",";
854 }
855 ss << ") is out of bound (";
856 for (int d = 0; d < N; ++d) {
857 ss << begin.vect[d] << ":" << end.vect[d]-1;
858 if (d + 1 < N) ss << ",";
859 };
860 ss << ")";
861 amrex::Abort(ss.str());
862 ))
863 }
864 }
865
866 // index_assert overload for M == N-1 last index assumed 0
867 // Only valid when last_dim_component == true
868 template <int M>
869 requires (((M+1 == N) || (M == AMREX_SPACEDIM)) && last_dim_component)
870#if defined(AMREX_USE_HIP)
872#else
874#endif
875 void index_assert (IntVectND<M> const& iv) const
876 {
877 IntVectND<N> iv_full = iv.template expand<N>(0);
878 for (int d = M; d < N; ++d) {
879 iv_full.vect[d] = begin.vect[d];
880 }
881 index_assert(iv_full);
882 }
883
884 // index_assert overload for M == N-1 and last index n
885 // Only valid when last_dim_component == true
886 template <int M>
887 requires (((M+1 == N) || (M == AMREX_SPACEDIM)) && last_dim_component)
888#if defined(AMREX_USE_HIP)
890#else
892#endif
893 void index_assert (IntVectND<M> const& iv, int n) const
894 {
895 IntVectND<N> iv_full = iv.template expand<N>(0);
896 for (int d = M; d < N-1; ++d) {
897 iv_full.vect[d] = begin.vect[d];
898 }
899 iv_full.vect[N-1] = n;
900 index_assert(iv_full);
901 }
902#endif
903
904 //
905 // Specialization for Array4
906 //
907
926 T& operator() (int i, int j, int k) const noexcept
927 requires (!std::is_void_v<T> && IsArray4_v)
928
929 {
930#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
931 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,0);
932#endif
933#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
934 return p[(i-begin.vect[0]) +
935 (j-begin.vect[1]) * stride.a[0] +
936 (k-begin.vect[2]) * stride.a[1]];
937#else
938 Long idx1 = i + j*stride.a[0] + k*stride.a[1];
939 Long idx0 = begin.vect[0]
940 + begin.vect[1] * stride.a[0]
941 + begin.vect[2] * stride.a[1];
942 return p[idx1-idx0];
943#endif
944 }
945
962 T& operator() (int i, int j, int k, int n) const noexcept
963 requires (!std::is_void_v<T> && IsArray4_v)
964
965 {
966#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
967 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,n);
968#endif
969#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
970 return p[(i-begin.vect[0]) +
971 (j-begin.vect[1]) * stride.a[0] +
972 (k-begin.vect[2]) * stride.a[1] +
973 n * stride.a[2]];
974#else
975 Long idx1 = i + j*stride.a[0] + k*stride.a[1] + n*stride.a[2];
976 Long idx0 = begin.vect[0]
977 + begin.vect[1] * stride.a[0]
978 + begin.vect[2] * stride.a[1];
979 return p[idx1-idx0];
980#endif
981 }
982
992 template <int M>
993 requires (!std::is_void_v<T> && IsArray4_v && (M == 3 || M == AMREX_SPACEDIM))
995 T& operator() (IntVectND<M> const& iv) const noexcept {
996#if (AMREX_SPACEDIM == 1)
997 if constexpr (M == 1) {
998 return this->operator()(iv.vect[0],0,0);
999 } else
1000#elif (AMREX_SPACEDIM == 2)
1001 if constexpr (M == 2) {
1002 return this->operator()(iv.vect[0],iv.vect[1],0);
1003 } else
1004#endif
1005 {
1006 return this->operator()(iv.vect[0],iv.vect[1],iv.vect[2]);
1007 }
1008 }
1009
1020 template <int M>
1021 requires (!std::is_void_v<T> && IsArray4_v && (M == 3 || M == AMREX_SPACEDIM))
1023 T& operator() (IntVectND<M> const& iv, int n) const noexcept {
1024#if (AMREX_SPACEDIM == 1)
1025 if constexpr (M == 1) {
1026 return this->operator()(iv.vect[0],0,0,n);
1027 } else
1028#elif (AMREX_SPACEDIM == 2)
1029 if constexpr (M == 2) {
1030 return this->operator()(iv.vect[0],iv.vect[1],0,n);
1031 } else
1032#endif
1033 {
1034 return this->operator()(iv.vect[0],iv.vect[1],iv.vect[2],n);
1035 }
1036 }
1037
1046 T& operator() (Dim3 const& cell) const noexcept
1047 requires (!std::is_void_v<T> && IsArray4_v)
1048
1049 {
1050 return this->operator()(cell.x,cell.y,cell.z);
1051 }
1052
1061 T& operator() (Dim3 const& cell, int n) const noexcept
1062 requires (!std::is_void_v<T> && IsArray4_v)
1063
1064 {
1065 return this->operator()(cell.x,cell.y,cell.z,n);
1066 }
1067
1077 T* ptr (int i, int j, int k) const noexcept
1078 requires (!std::is_void_v<T> && IsArray4_v)
1079
1080 {
1081#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
1082 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,0);
1083#endif
1084#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
1085 return p + ((i-begin.vect[0]) +
1086 (j-begin.vect[1]) * stride.a[0] +
1087 (k-begin.vect[2]) * stride.a[1]);
1088#else
1089 Long idx1 = i + j*stride.a[0] + k*stride.a[1];
1090 Long idx0 = begin.vect[0]
1091 + begin.vect[1] * stride.a[0]
1092 + begin.vect[2] * stride.a[1];
1093 return p + (idx1-idx0);
1094#endif
1095 }
1096
1107 T* ptr (int i, int j, int k, int n) const noexcept
1108 requires (!std::is_void_v<T> && IsArray4_v)
1109
1110 {
1111#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
1112 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,n);
1113#endif
1114#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
1115 return p + ((i-begin.vect[0]) +
1116 (j-begin.vect[1]) * stride.a[0] +
1117 (k-begin.vect[2]) * stride.a[1] +
1118 n * stride.a[2]);
1119#else
1120 Long idx1 = i + j*stride.a[0] + k*stride.a[1] + n*stride.a[2];
1121 Long idx0 = begin.vect[0]
1122 + begin.vect[1] * stride.a[0]
1123 + begin.vect[2] * stride.a[1];
1124 return p + (idx1-idx0);
1125#endif
1126 }
1127
1134 template <int M>
1135 requires (!std::is_void_v<T> && IsArray4_v && (M == 3 || M == AMREX_SPACEDIM))
1137 T* ptr (IntVectND<M> const& iv) const noexcept {
1138#if (AMREX_SPACEDIM == 1)
1139 if constexpr (M == 1) {
1140 return this->ptr(iv.vect[0],0,0);
1141 } else
1142#elif (AMREX_SPACEDIM == 2)
1143 if constexpr (M == 2) {
1144 return this->ptr(iv.vect[0],iv.vect[1],0);
1145 } else
1146#endif
1147 {
1148 return this->ptr(iv.vect[0],iv.vect[1],iv.vect[2]);
1149 }
1150 }
1151
1159 template <int M>
1160 requires (!std::is_void_v<T> && IsArray4_v && (M == 3 || M == AMREX_SPACEDIM))
1162 T* ptr (IntVectND<M> const& iv, int n) const noexcept {
1163#if (AMREX_SPACEDIM == 1)
1164 if constexpr (M == 1) {
1165 return this->ptr(iv.vect[0],0,0,n);
1166 } else
1167#elif (AMREX_SPACEDIM == 2)
1168 if constexpr (M == 2) {
1169 return this->ptr(iv.vect[0],iv.vect[1],0,n);
1170 } else
1171#endif
1172 {
1173 return this->ptr(iv.vect[0],iv.vect[1],iv.vect[2],n);
1174 }
1175 }
1176
1184 T* ptr (Dim3 const& cell) const noexcept
1185 requires (!std::is_void_v<T> && IsArray4_v)
1186
1187 {
1188 return this->ptr(cell.x,cell.y,cell.z);
1189 }
1190
1199 T* ptr (Dim3 const& cell, int n) const noexcept
1200 requires (!std::is_void_v<T> && IsArray4_v)
1201
1202 {
1203 return this->ptr(cell.x,cell.y,cell.z,n);
1204 }
1205
1215 bool contains (int i, int j, int k) const noexcept
1216 requires (IsArray4_v)
1217
1218 {
1219 return (i>=begin.vect[0] && i<end.vect[0] &&
1220 j>=begin.vect[1] && j<end.vect[1] &&
1221 k>=begin.vect[2] && k<end.vect[2]);
1222 }
1223
1230 template <int M>
1231 requires (IsArray4_v && (M == 3 || M == AMREX_SPACEDIM))
1233 bool contains (IntVectND<M> const& iv) const noexcept {
1234#if (AMREX_SPACEDIM < 3)
1235 if constexpr (M == AMREX_SPACEDIM) {
1236 return AMREX_D_TERM((iv.vect[0]>=begin.vect[0]) && (iv.vect[0]<end.vect[0]),
1237 && (iv.vect[1]>=begin.vect[1]) && (iv.vect[1]<end.vect[1]),
1238 && (iv.vect[2]>=begin.vect[2]) && (iv.vect[2]<end.vect[2]));
1239 } else
1240#endif
1241 {
1242 return (iv.vect[0]>=begin.vect[0]) && (iv.vect[0]<end.vect[0])
1243 && (iv.vect[1]>=begin.vect[1]) && (iv.vect[1]<end.vect[1])
1244 && (iv.vect[2]>=begin.vect[2]) && (iv.vect[2]<end.vect[2]);
1245 }
1246 }
1247
1255 bool contains (Dim3 const& cell) const noexcept
1256 requires (IsArray4_v)
1257
1258 {
1259 return (cell.x>=begin.vect[0] && cell.x<end.vect[0] &&
1260 cell.y>=begin.vect[1] && cell.y<end.vect[1] &&
1261 cell.z>=begin.vect[2] && cell.z<end.vect[2]);
1262 }
1263
1264#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
1265#if defined(AMREX_USE_HIP)
1267#else
1269#endif
1270 void index_assert_print_error_message (int i, int j, int k, int n) const
1271 requires (IsArray4_v)
1272
1273 {
1275 AMREX_DEVICE_PRINTF(" (%d,%d,%d,%d) is out of bound (%d:%d,%d:%d,%d:%d,0:%d)\n",
1276 i, j, k, n,
1277 begin.vect[0], end.vect[0]-1,
1278 begin.vect[1], end.vect[1]-1,
1279 begin.vect[2], end.vect[2]-1,
1280 end.vect[3]-1);
1281 amrex::Abort();
1282 ))
1284 std::stringstream ss;
1285 ss << " (" << i << "," << j << "," << k << "," << n
1286 << ") is out of bound ("
1287 << begin.vect[0] << ":" << end.vect[0]-1 << ","
1288 << begin.vect[1] << ":" << end.vect[1]-1 << ","
1289 << begin.vect[2] << ":" << end.vect[2]-1 << ","
1290 << "0:" << end.vect[3]-1 << ")";
1291 amrex::Abort(ss.str());
1292 ))
1293 }
1294#endif
1295
1296 private:
1298 constexpr void set_stride () noexcept {
1299 if constexpr (N > 1) {
1300 Long current_stride = 1;
1301 constexpr_for<0, N-1>([&](int d) {
1302 Long len = end.vect[d] - begin.vect[d];
1303 current_stride *= len;
1304 stride.a[d] = current_stride;
1305 });
1306 }
1307 }
1308 };
1309
1311 // Deduction guides for ArrayND
1312 // 1: Matches ArrayND (T*, BoxND<N>) -> ArrayND<T,N, last_dim_component=false>
1313 template <typename T, int N>
1314 ArrayND (T*, BoxND<N> const&) -> ArrayND<T, N, false>;
1315
1316 // 2: Matches ArrayND (T*, BoxND<N>, ncomp) -> ArrayND<T,N+1, last_dim_component=true>
1317 // This supports the "N+1" logic (Spatial + Component)
1318 template <typename T, int N>
1319 ArrayND (T*, BoxND<N> const&, int) -> ArrayND<T, N+1, true>;
1320
1321 // 3: Matches ArrayND (T*, IntVectND<N>, IntVectND<N>) -> ArrayND<T,N, last_dim_component=false>
1322 template <typename T, int N>
1323 ArrayND (T*, IntVectND<N> const&, IntVectND<N> const&) -> ArrayND<T, N, false>;
1324
1325 // 4: Matches ArrayND (T*, IntVectND<N>, IntVectND<N>, int) -> ArrayND<T,N+1, last_dim_component=true>
1326 template <typename T, int N>
1327 ArrayND (T*, IntVectND<N> const&, IntVectND<N> const&, int) -> ArrayND<T, N+1, true>;
1328
1329 // 5: Matches ArrayND (T*, Dim3, Dim3) -> ArrayND<T,4, last_dim_component=true>
1330 template <typename T>
1331 ArrayND (T*, Dim3 const&, Dim3 const&, int) -> ArrayND<T, 4, true>;
1333
1339 template<typename T>
1341
1349 template <class T>
1351 Dim3 lbound (Array4<T> const& a) noexcept
1352 {
1353 return Dim3{.x = a.begin.vect[0], .y = a.begin.vect[1], .z = a.begin.vect[2]};
1354 }
1355
1363 template <class T>
1365 Dim3 ubound (Array4<T> const& a) noexcept
1366 {
1367 return Dim3{.x = a.end.vect[0]-1, .y = a.end.vect[1]-1, .z = a.end.vect[2]-1};
1368 }
1369
1377 template <class T>
1379 Dim3 length (Array4<T> const& a) noexcept
1380 {
1381 return Dim3{.x = a.end.vect[0]-a.begin.vect[0],
1382 .y = a.end.vect[1]-a.begin.vect[1],
1383 .z = a.end.vect[2]-a.begin.vect[2]};
1384 }
1385
1397 template <typename T, int N, bool C>
1398 std::ostream& operator<< (std::ostream& os, const ArrayND<T,N,C>& a) {
1399 os << "(" << a.begin << ',' << a.end-1 << ")";
1400 return os;
1401 }
1402
1404 //
1405 // Type traits for detecting if a class has a size() constexpr function.
1406 //
1407 template <class A> struct HasMultiComp : std::false_type {};
1408 //
1409 template <class B>
1410 requires (B().size() >= 1)
1411 struct HasMultiComp<B> : std::true_type {};
1413
1426 template <typename T>
1428 : public Array4<T>
1429 {
1432 : Array4<T>{a} {}
1433
1435 T& operator() (int i, int j, int k) const noexcept {
1436 return this->Array4<T>::operator()(i,j,k);
1437 }
1438
1440 typename T::reference_type
1441 operator() (int i, int j, int k, int n) const noexcept
1442 requires (amrex::HasMultiComp<T>::value)
1443
1444 {
1445 return this->Array4<T>::operator()(i,j,k,0)[n];
1446 }
1447
1449 T& operator() (int i, int j, int k, int n) const noexcept
1450 requires (!amrex::HasMultiComp<T>::value)
1451
1452 {
1453 return this->Array4<T>::operator()(i,j,k,n);
1454 }
1455 };
1456
1459 template <typename T>
1460 [[nodiscard]] PolymorphicArray4<T>
1462 {
1463 return PolymorphicArray4<T>(a);
1464 }
1465}
1466
1467#endif
Runtime initialization/finalization helpers and global diagnostics.
Compile-time unrolled loop utility.
#define AMREX_NO_UNIQUE_ADDRESS
Definition AMReX_Extension.H:307
#define AMREX_NO_INLINE
Definition AMReX_Extension.H:141
#define AMREX_FORCE_INLINE
Definition AMReX_Extension.H:124
#define AMREX_RESTRICT
Definition AMReX_Extension.H:37
#define AMREX_DEVICE_PRINTF(...)
Definition AMReX_GpuPrint.H:15
#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
Array4< int const > offset
Definition AMReX_HypreMLABecLap.cpp:1131
#define AMREX_D_TERM(a, b, c)
Definition AMReX_SPACE.H:172
A Rectangular Domain on an Integer Lattice.
Definition AMReX_Box.H:54
An Integer Vector in dim-Dimensional Space.
Definition AMReX_IntVect.H:149
__host__ __device__ constexpr bool allGT(const IntVectND< dim > &rhs) const noexcept
Returns true if this is greater than argument for all components. NOTE: This is NOT a strict weak ord...
Definition AMReX_IntVect.H:517
int vect[dim]
Definition AMReX_IntVect.H:885
amrex_long Long
Definition AMReX_INT.H:30
__host__ __device__ Dim3 ubound(Array4< T > const &a) noexcept
Return the inclusive upper bounds of an Array4 in Dim3 form.
Definition AMReX_Array4.H:1365
__host__ __device__ Dim3 length(Array4< T > const &a) noexcept
Return the spatial extents of an Array4 in Dim3 form.
Definition AMReX_Array4.H:1379
PolymorphicArray4< T > makePolymorphic(Array4< T > const &a)
Definition AMReX_Array4.H:1461
__host__ __device__ Dim3 lbound(Array4< T > const &a) noexcept
Return the inclusive lower bounds of an Array4 in Dim3 form.
Definition AMReX_Array4.H:1351
__host__ __device__ constexpr void constexpr_for(F const &f)
Compile-time unrolled loop from I (inclusive) to N (exclusive).
Definition AMReX_ConstexprFor.H:38
Definition AMReX_Amr.cpp:50
std::ostream & operator<<(std::ostream &os, AmrMesh const &amr_mesh)
Stream helper; forwards to the friend declared inside AmrMesh.
Definition AMReX_AmrMesh.cpp:1372
__host__ __device__ Dim3 begin(BoxND< dim > const &box) noexcept
Return the iterator begin coordinate of box as Dim3.
Definition AMReX_Box.H:2239
void Abort(const std::string &msg)
Print a fatal-error message to stderr and abort execution.
Definition AMReX.cpp:242
__host__ __device__ Dim3 end(BoxND< dim > const &box) noexcept
Return the iterator end coordinate of box as Dim3.
Definition AMReX_Box.H:2257
A multidimensional array accessor.
Definition AMReX_Array4.H:289
__host__ __device__ T * ptr(int i, int j, int k) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1077
__host__ __device__ constexpr bool contains(IntVectND< M > const &iv, int n) const noexcept
Test whether a spatial index and component lie inside the bounds.
Definition AMReX_Array4.H:742
static constexpr bool IsArray4_v
True if this is an Array4 (N==4 and last dim is component).
Definition AMReX_Array4.H:296
__host__ __device__ ArrayND(T *a_p, BoxND< M > const &box, int ncomp) noexcept
Constructor using a BoxND and the number of components.
Definition AMReX_Array4.H:350
__host__ __device__ constexpr bool ok() const noexcept
Check if the ArrayND pointer is valid and bounds are valid.
Definition AMReX_Array4.H:476
__host__ __device__ T & operator()(idx... i) const noexcept
Multi-index operator() for accessing elements.
Definition AMReX_Array4.H:498
__host__ __device__ ArrayND(T *a_p, BoxND< N > const &box) noexcept
Constructor using a BoxND.
Definition AMReX_Array4.H:334
__host__ __device__ constexpr std::size_t size() const noexcept
Total number of elements in the ArrayND's index region.
Definition AMReX_Array4.H:643
__host__ __device__ constexpr bool contains(IntVectND< M > const &iv) const noexcept
Test whether an IntVectND lies inside the ArrayND bounds.
Definition AMReX_Array4.H:715
__host__ __device__ bool contains(IntVectND< M > const &iv) const noexcept
Test whether the spatial indices are inside the Array4 bounds.
Definition AMReX_Array4.H:1233
__host__ __device__ T * ptr(IntVectND< M > const &iv) const noexcept
Access pointer by IntVectND.
Definition AMReX_Array4.H:584
T *__restrict__ p
Definition AMReX_Array4.H:298
__host__ __device__ constexpr ArrayND(T *a_p, IntVectND< N > const &a_begin, IntVectND< N > const &a_end) noexcept
IntVectND<N> constructor.
Definition AMReX_Array4.H:364
__host__ __device__ constexpr bool contains(idx... i) const noexcept
Test whether an index tuple lies inside the ArrayND bounds.
Definition AMReX_Array4.H:691
__host__ __device__ T * ptr(Dim3 const &cell, int n) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1199
static constexpr bool IsLastDimComponent_v
True if the last dimension is treated as components.
Definition AMReX_Array4.H:294
__host__ __device__ constexpr int nComp() const noexcept
Get number of components.
Definition AMReX_Array4.H:626
__host__ __device__ constexpr T * dataPtr() const noexcept
Get raw data pointer.
Definition AMReX_Array4.H:617
__host__ __device__ T * ptr(idx... i) const noexcept
Multi-index ptr() for accessing pointer to element.
Definition AMReX_Array4.H:561
__host__ __device__ T * ptr(int i, int j, int k, int n) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1107
IntVectND< N > end
Exclusive upper bounds.
Definition AMReX_Array4.H:303
__host__ __device__ constexpr Long get_offset(IntVectND< M > const &iv, int n) const noexcept
Compute the linear offset (in elements) for an IntVectND and component index.
Definition AMReX_Array4.H:819
IntVectND< N > begin
Inclusive lower bounds.
Definition AMReX_Array4.H:302
__host__ __device__ CellData< T > cellData(int i, int j, int k) const noexcept
Create a single-cell component accessor.
Definition AMReX_Array4.H:768
__host__ __device__ T * ptr(IntVectND< M > const &iv, int n) const noexcept
Access pointer by spatial IntVectND and component index.
Definition AMReX_Array4.H:605
__host__ __device__ T * ptr(IntVectND< M > const &iv) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1137
__host__ __device__ constexpr ArrayND(ArrayND< U, N, last_dim_component > const &rhs, int start_comp, int num_comp) noexcept
Slicing constructor (Component subset with count).
Definition AMReX_Array4.H:453
__host__ __device__ constexpr Long get_offset(IntVectND< M > const &iv) const noexcept
Compute the linear offset (in elements) for an IntVectND.
Definition AMReX_Array4.H:786
__host__ __device__ T * ptr(Dim3 const &cell) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1184
__host__ __device__ constexpr Long get_stride() const noexcept
Return the stride (in elements) for dimension d.
Definition AMReX_Array4.H:668
__host__ __device__ constexpr ArrayND() noexcept
Default-construct an empty accessor.
Definition AMReX_Array4.H:311
__host__ __device__ constexpr ArrayND(ArrayND< U, N, last_dim_component > const &rhs, int start_comp) noexcept
Slicing constructor (Component subset).
Definition AMReX_Array4.H:430
__host__ __device__ constexpr ArrayND(T *a_p, Dim3 const &a_begin, Dim3 const &a_end, int a_ncomp) noexcept
Constructor for N=4 using Dim3.
Definition AMReX_Array4.H:382
__host__ __device__ T * ptr(IntVectND< M > const &iv, int n) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1162
__host__ __device__ bool contains(Dim3 const &cell) const noexcept
Test whether the spatial indices are inside the Array4 bounds.
Definition AMReX_Array4.H:1255
__host__ __device__ bool contains(int i, int j, int k) const noexcept
Test whether the spatial indices are inside the Array4 bounds.
Definition AMReX_Array4.H:1215
__host__ __device__ constexpr ArrayND(T *a_p, IntVectND< M > const &a_begin, IntVectND< M > const &a_end, int ncomp) noexcept
Reduced dimension constructor with component count.
Definition AMReX_Array4.H:402
Lightweight accessor for data associated with a single cell.
Definition AMReX_Array4.H:47
__host__ __device__ constexpr CellData(T *a_p, Long a_stride, int a_ncomp)
Construct a CellData.
Definition AMReX_Array4.H:64
__host__ __device__ T & operator[](int n) const
Access the n-th component of the cell.
Definition AMReX_Array4.H:106
__host__ __device__ int nComp() const noexcept
Return the number of components referenced by this accessor.
Definition AMReX_Array4.H:93
A simple struct holding 3 int values for a 3D index.
Definition AMReX_Dim3.H:24
int x
Definition AMReX_Dim3.H:24
Array4 subclass that provides uniform (i,j,k,n) access for both AoS and SoA data.
Definition AMReX_Array4.H:1429
__host__ __device__ PolymorphicArray4(Array4< T > const &a)
Definition AMReX_Array4.H:1431