4#include <AMReX_Config.H>
15#define AMREX_ARRAY4_INDEX_ASSERT(i,j,k,n) \
16 if ((i)<begin.vect[0] || (i)>=end.vect[0] || \
17 (j)<begin.vect[1] || (j)>=end.vect[1] || \
18 (k)<begin.vect[2] || (k)>=end.vect[2] || \
19 (n)<0 || (n)>=end.vect[3]) \
21 index_assert_print_error_message(i,j,k,n); \
60 : p(a_p), stride(a_stride), ncomp(a_ncomp)
72 std::enable_if_t<std::is_const_v<U>,
int> = 0>
75 : p(rhs.p), stride(rhs.stride), ncomp(rhs.ncomp)
84 explicit operator bool() const noexcept {
return p !=
nullptr; }
90 int nComp() const noexcept {
return ncomp; }
103 std::enable_if_t<!std::is_void_v<U>,
int> = 0>
106#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
107 if (n < 0 || n >= ncomp) {
112 std::stringstream ss;
113 ss <<
" " << n <<
" is out of bound: (0:" << ncomp-1 <<
")";
125 template <
int N>
struct Stride { Long a[N] = {}; };
126 template <>
struct Stride<0> {};
133 template <
typename T>
134 struct IsValidIndexType {
135 static constexpr bool value = std::is_integral_v<T> && IsNonNarrowingConversion_v<T, int>;
144 template <
int N,
bool last_dim_component,
typename... idx>
145 struct ArrayNDIndexCheck_impl {
147 static constexpr int num_indices =
sizeof...(idx);
149 template <
typename... Ts>
150 struct AllExceptLastEnum;
152 template <
typename first,
typename... Ts>
153 struct AllExceptLastEnum<first, Ts...> {
154 static constexpr bool value =
155 IsValidIndexType<std::decay_t<first>>::value
156 && AllExceptLastEnum<Ts...>::value;
161 template <
typename last>
162 struct AllExceptLastEnum<last> {
163 static constexpr bool value = IsValidIndexType<std::decay_t<last>>::value
164 || (std::is_enum_v<std::decay_t<last>>);
167 static constexpr bool index_with_maybe_enum = AllExceptLastEnum<idx...>::value;
169 static constexpr bool index_all_values = Conjunction<
170 IsValidIndexType<std::decay_t<idx>>...>::value;
173 static constexpr bool value = ((num_indices == N) && index_all_values)
174 || ((num_indices == N - 1) && index_all_values && last_dim_component)
175 || ((num_indices == N) && index_with_maybe_enum && last_dim_component);
178 template <
int N,
bool last_dim_component,
class... idx>
179 inline constexpr bool ArrayNDIndexCheck_impl_v = ArrayNDIndexCheck_impl<N, last_dim_component, idx...>::value;
181 template<std::size_t... idx>
182 constexpr auto make_oob_message_impl (std::index_sequence<idx...>) {
183 constexpr std::size_t N =
sizeof...(idx);
184 constexpr char prefix[] =
" (";
185 constexpr char middle[] =
") is out of bound (";
186 constexpr char suffix[] =
")\n";
188 constexpr std::size_t size =
195 std::array<char, size> buf{};
200 for (
char c : prefix) {
206 ((buf[pos++] =
'%', buf[pos++] =
'd', idx + 1 < N ? buf[pos++] =
',' : 0), ...);
209 for (
char c : middle) {
216 ((buf[pos++] =
'%', buf[pos++] =
'd', buf[pos++] =
':', buf[pos++] =
'%', buf[pos++] =
'd',
217 idx + 1 < N ? buf[pos++] =
',' : 0), ...);
220 for (
char c : suffix) {
226 template <std::size_t... idx, std::size_t... idx2x>
228 void device_print_impl2 (std::index_sequence<idx...>,
229 std::index_sequence<idx2x...>,
230 IntVectND<
sizeof...(idx)>
const& iv,
231 IntVectND<
sizeof...(idx)>
const&
begin,
232 IntVectND<
sizeof...(idx)>
const&
end)
234 constexpr auto msg = make_oob_message_impl(std::index_sequence<idx...>{});
238 ((idx2x % 2 == 0) ?
begin.vect[idx2x / 2] :
end.vect[idx2x / 2])...
242 template <
int N, std::enable_if_t<(N>=1),
int> = 0>
244 void device_printf_impl (
const IntVectND<N>& iv,
245 const IntVectND<N>&
begin,
246 const IntVectND<N>&
end)
249 std::make_index_sequence<N>{},
250 std::make_index_sequence<N*2>{},
281 template<
typename T,
int N,
bool last_dim_component = false>
284 static_assert(N >= 1,
"ArrayND must have at least one dimension");
285 static_assert(N > 1 || !last_dim_component,
"ArrayND with N=1 cannot have last_dim_component=true");
290 static constexpr bool IsArray4_v = (N==4 && last_dim_component);
315 template <
class U=T, std::enable_if_t<std::is_const_v<U>,
int> = 0>
317 constexpr ArrayND (
ArrayND<std::remove_const_t<T>, N, last_dim_component>
const& rhs) noexcept
318 :
p(rhs.p), stride(rhs.stride),
begin(rhs.begin),
end(rhs.end) {}
325 template <
bool C = last_dim_component, std::enable_if_t<!C,
int> = 0>
329 :
ArrayND(a_p, box.smallEnd(), box.bigEnd() + 1)
338 template <
int M, std::enable_if_t<((M+1==N) || (N == 4 && M == AMREX_SPACEDIM))
339 && last_dim_component,
int> = 0>
355 template <
bool C = last_dim_component, std::enable_if_t<!C,
int> = 0>
373 template <
bool B = IsArray4_v, std::enable_if_t<B,
int> = 0>
375 constexpr ArrayND (T* a_p,
Dim3 const& a_begin,
Dim3 const& a_end,
int a_ncomp) noexcept
376 :
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)
391 template <
int M, std::enable_if_t<((M+1 == N) || (N == 4 && M == AMREX_SPACEDIM))
392 && last_dim_component,
int> = 0>
397 constexpr_for<0, M>([&](
int d) {
401 constexpr_for<M, N>([&](
int d) {
420 <std::is_same_v<std::remove_const_t<T>, std::remove_const_t<U>>
421 && (N >= 2) && last_dim_component,
int> = 0>
424 :
p((T*)(rhs.
p + start_comp*rhs.stride.a[N-2])),
444 <std::is_same_v<std::remove_const_t<T>, std::remove_const_t<U>>
445 && (N >= 2) && last_dim_component,
int> = 0>
448 :
p((T*)(rhs.
p + start_comp*rhs.stride.a[N-2])),
462 constexpr explicit operator bool() const noexcept {
return p !=
nullptr; }
488 template <
typename... idx,
class U=T,
489 std::enable_if_t<!std::is_void_v<U> && !
IsArray4_v
490 && detail::ArrayNDIndexCheck_impl_v<N, last_dim_component, idx...>,
int> = 0>
494 constexpr auto nidx =
sizeof...(i);
495#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
513 template <
int M,
class U=T, std::enable_if_t<
514 !std::is_void_v<U> &&
515 ((M == N) || (!
IsArray4_v && last_dim_component && (M + 1 == N))),
int> = 0>
518#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
535 template <
int M,
class U=T, std::enable_if_t<
536 !std::is_void_v<U> && last_dim_component && !
IsArray4_v
537 && (M + 1 == N),
int> = 0>
540#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
552 template <
typename... idx,
class U=T,
553 std::enable_if_t<!std::is_void_v<U> && !
IsArray4_v
554 && detail::ArrayNDIndexCheck_impl_v<N, last_dim_component, idx...>,
int> = 0>
556 T*
ptr (idx... i)
const noexcept {
557 constexpr auto nidx =
sizeof...(i);
558#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
576 template <
int M, std::enable_if_t<(M == N)
577 || (!
IsArray4_v && last_dim_component && (M + 1 == N)),
int> = 0>
580#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
597 template <
int M, std::enable_if_t<last_dim_component && !
IsArray4_v
598 && (M + 1 == N),
int> = 0>
601#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
621 constexpr int nComp () const noexcept {
622 if constexpr (last_dim_component) {
638 constexpr std::size_t
size () const noexcept {
641 constexpr_for<0, N>([&](
int d) {
660 template <
int d, std::enable_if_t<(d < N) && (d >= 0),
int> = 0>
663 if constexpr (N > 1 && d > 0) {
664 return stride.a[d-1];
681 template <
typename... idx, std::enable_if_t<!
IsArray4_v &&
682 detail::ArrayNDIndexCheck_impl_v<N, last_dim_component, idx...>,
int> = 0>
684 constexpr bool contains (idx... i)
const noexcept {
685 constexpr auto nidx =
sizeof...(i);
705 template <
int M, std::enable_if_t<(M == N)
706 || (!
IsArray4_v && last_dim_component && (M + 1 == N)),
int> = 0>
710 constexpr_for<0, M>([&](
int d) {
732 template <
int M, std::enable_if_t<last_dim_component && !
IsArray4_v
733 && (M + 1 == N),
int> = 0>
737 constexpr_for<0, M>([&](
int d) {
740 inside = inside && (n >= 0) && (n <
end.
vect[N-1]);
760 template <
int M=N,
bool B = IsArray4_v, std::enable_if_t<B,
int> = 0>
774 template <
int M, std::enable_if_t<(M == N)
775 || (last_dim_component && (M + 1 == N || M == AMREX_SPACEDIM)),
int> = 0>
782 constexpr int idx = (last_dim_component && M == N) ? N - 1 : M;
783 if constexpr (N > 1) {
784 constexpr_for<1, idx>([&](
int d) {
791 if constexpr (last_dim_component && M == N) {
806 template <
int M, std::enable_if_t<last_dim_component
807 && ((M + 1 == N) || (N == 4 && M == AMREX_SPACEDIM)),
int> = 0>
812 constexpr_for<1, M>([&](
int d) {
815 offset += n * stride.a[N-2];
819#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
820#if defined(AMREX_USE_HIP)
827 bool out_of_bounds =
false;
828 for (
int d = 0; d < N; ++d) {
830 out_of_bounds =
true;
835 detail::device_printf_impl(iv,
begin,
end);
839 std::stringstream ss;
841 for (
int d = 0; d < N; ++d) {
843 if (d + 1 < N) ss <<
",";
845 ss <<
") is out of bound (";
846 for (
int d = 0; d < N; ++d) {
848 if (d + 1 < N) ss <<
",";
858 template <
int M, std::enable_if_t<((M+1 == N) || (M == AMREX_SPACEDIM))
859 && last_dim_component,
int> = 0>
860#
if defined(AMREX_USE_HIP)
865 void index_assert (IntVectND<M>
const& iv)
const
867 IntVectND<N> iv_full = iv.template expand<N>(0);
868 for (
int d = M; d < N; ++d) {
871 index_assert(iv_full);
876 template <
int M, std::enable_if_t<((M+1 == N) || (M == AMREX_SPACEDIM))
877 && last_dim_component,
int> = 0>
878#
if defined(AMREX_USE_HIP)
883 void index_assert (IntVectND<M>
const& iv,
int n)
const
885 IntVectND<N> iv_full = iv.template expand<N>(0);
886 for (
int d = M; d < N-1; ++d) {
889 iv_full.vect[N-1] = n;
890 index_assert(iv_full);
915 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
918#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
919 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,0);
921#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
926 Long idx1 = i + j*stride.a[0] + k*stride.a[1];
949 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
952#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
953 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,n);
955#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
961 Long idx1 = i + j*stride.a[0] + k*stride.a[1] + n*stride.a[2];
978 template <
int M,
class U=T,
979 std::enable_if_t<!std::is_void_v<U> &&
IsArray4_v && (M == 3 || M == AMREX_SPACEDIM),
int> = 0>
982#if (AMREX_SPACEDIM == 1)
983 if constexpr (M == 1) {
986#elif (AMREX_SPACEDIM == 2)
987 if constexpr (M == 2) {
988 return this->
operator()(iv.vect[0],iv.vect[1],0);
992 return this->
operator()(iv.vect[0],iv.vect[1],iv.vect[2]);
1006 template <
int M,
class U=T,
1007 std::enable_if_t<!std::is_void_v<U> &&
IsArray4_v && (M == 3 || M == AMREX_SPACEDIM),
int> = 0>
1010#if (AMREX_SPACEDIM == 1)
1011 if constexpr (M == 1) {
1014#elif (AMREX_SPACEDIM == 2)
1015 if constexpr (M == 2) {
1031 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
1034 return this->
operator()(cell.x,cell.y,cell.z);
1044 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
1047 return this->
operator()(cell.x,cell.y,cell.z,n);
1058 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
1060 U*
ptr (
int i,
int j,
int k)
const noexcept {
1061#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
1062 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,0);
1064#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
1069 Long idx1 = i + j*stride.a[0] + k*stride.a[1];
1073 return p + (idx1-idx0);
1086 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
1088 U*
ptr (
int i,
int j,
int k,
int n)
const noexcept {
1089#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
1090 AMREX_ARRAY4_INDEX_ASSERT(i,j,k,n);
1092#if defined(AMREX_USE_GPU) || defined(AMREX_DEBUG)
1098 Long idx1 = i + j*stride.a[0] + k*stride.a[1] + n*stride.a[2];
1102 return p + (idx1-idx0);
1112 template <
int M,
class U=T,
1113 std::enable_if_t<!std::is_void_v<U> &&
IsArray4_v && (M == 3 || M == AMREX_SPACEDIM),
int> = 0>
1116#if (AMREX_SPACEDIM == 1)
1117 if constexpr (M == 1) {
1118 return this->
ptr(iv.
vect[0],0,0);
1120#elif (AMREX_SPACEDIM == 2)
1121 if constexpr (M == 2) {
1137 template <
int M,
class U=T,
1138 std::enable_if_t<!std::is_void_v<U> &&
IsArray4_v && (M == 3 || M == AMREX_SPACEDIM),
int> = 0>
1141#if (AMREX_SPACEDIM == 1)
1142 if constexpr (M == 1) {
1143 return this->
ptr(iv.
vect[0],0,0,n);
1145#elif (AMREX_SPACEDIM == 2)
1146 if constexpr (M == 2) {
1161 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
1164 return this->
ptr(cell.x,cell.y,cell.z);
1174 template <
class U=T, std::enable_if_t<!std::is_
void_v<U> && IsArray4_v,
int> = 0>
1177 return this->
ptr(cell.x,cell.y,cell.z,n);
1188 template <
bool B = IsArray4_v, std::enable_if_t<B,
int> = 0>
1202 template <
int M, std::enable_if_t<IsArray4_v && (M == 3 || M == AMREX_SPACEDIM),
int> = 0>
1205#if (AMREX_SPACEDIM < 3)
1206 if constexpr (M == AMREX_SPACEDIM) {
1225 template <
bool B = IsArray4_v, std::enable_if_t<B,
int> = 0>
1233#if defined(AMREX_DEBUG) || defined(AMREX_BOUND_CHECK)
1234 template <
bool B = IsArray4_v, std::enable_if_t<B,
int> = 0>
1235#if defined(AMREX_USE_HIP)
1240 void index_assert_print_error_message (
int i,
int j,
int k,
int n)
const
1252 std::stringstream ss;
1253 ss << " (" << i << "," << j << "," << k << "," << n
1254 << ") is out of bound ("
1255 <<
begin.vect[0] << ":" <<
end.vect[0]-1 << ","
1256 <<
begin.vect[1] << ":" <<
end.vect[1]-1 << ","
1257 <<
begin.vect[2] << ":" <<
end.vect[2]-1 << ","
1258 << "0:" <<
end.vect[3]-1 << ")";
1266 constexpr void set_stride () noexcept {
1267 if constexpr (N > 1) {
1268 Long current_stride = 1;
1271 current_stride *= len;
1272 stride.a[d] = current_stride;
1280 template <
typename T,
int N>
1285 template <
typename T,
int N>
1289 template <
typename T,
int N>
1293 template <
typename T,
int N>
1297 template <
typename T>
1305 template<
typename T>
1319 return Dim3{a.begin.vect[0],a.begin.vect[1],a.begin.vect[2]};
1333 return Dim3{a.end.vect[0]-1,a.end.vect[1]-1,a.end.vect[2]-1};
1347 return Dim3{a.end.vect[0]-a.begin.vect[0],a.end.vect[1]-a.begin.vect[1],a.end.vect[2]-a.begin.vect[2]};
1361 template <
typename T,
int N,
bool C>
1363 os <<
"(" << a.
begin <<
',' << a.
end-1 <<
")";
1370 template <
class A,
class Enable =
void>
struct HasMultiComp : std::false_type {};
1374 : std::true_type {};
1381 template <
typename T>
1390 T& operator() (
int i,
int j,
int k)
const noexcept {
1394 template <class U=T, std::enable_if_t< amrex::HasMultiComp<U>::value,
int> = 0>
1396 typename U::reference_type
1397 operator() (
int i,
int j,
int k,
int n)
const noexcept {
1401 template <class U=T, std::enable_if_t<!amrex::HasMultiComp<U>::value,
int> = 0>
1403 U& operator() (
int i,
int j,
int k,
int n)
const noexcept {
1408 template <
typename T>
1409 [[nodiscard]] PolymorphicArray4<T>
#define AMREX_NO_UNIQUE_ADDRESS
Definition AMReX_Extension.H:263
#define AMREX_NO_INLINE
Definition AMReX_Extension.H:136
#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
Array4< int const > offset
Definition AMReX_HypreMLABecLap.cpp:1089
#define AMREX_D_TERM(a, b, c)
Definition AMReX_SPACE.H:172
A Rectangular Domain on an Integer Lattice.
Definition AMReX_Box.H:49
__host__ __device__ const IntVectND< dim > & bigEnd() const &noexcept
Return the inclusive upper bound of the box.
Definition AMReX_Box.H:123
__host__ __device__ const IntVectND< dim > & smallEnd() const &noexcept
Return the inclusive lower bound of the box.
Definition AMReX_Box.H:111
An Integer Vector in dim-Dimensional Space.
Definition AMReX_IntVect.H:57
__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:425
int vect[dim]
Definition AMReX_IntVect.H:793
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:1331
__host__ __device__ Dim3 length(Array4< T > const &a) noexcept
Return the spatial extents of an Array4 in Dim3 form.
Definition AMReX_Array4.H:1345
__host__ __device__ Dim3 lbound(Array4< T > const &a) noexcept
Return the inclusive lower bounds of an Array4 in Dim3 form.
Definition AMReX_Array4.H:1317
Definition AMReX_Amr.cpp:49
__host__ __device__ Dim3 begin(BoxND< dim > const &box) noexcept
Definition AMReX_Box.H:2006
__host__ __device__ constexpr void constexpr_for(F const &f)
Definition AMReX_ConstexprFor.H:28
PolymorphicArray4< T > makePolymorphic(Array4< T > const &a)
Definition AMReX_Array4.H:1410
void Abort(const std::string &msg)
Print out message to cerr and exit via abort().
Definition AMReX.cpp:230
__host__ __device__ Dim3 end(BoxND< dim > const &box) noexcept
Definition AMReX_Box.H:2015
std::ostream & operator<<(std::ostream &os, AmrMesh const &amr_mesh)
Definition AMReX_AmrMesh.cpp:1237
A multidimensional array accessor.
Definition AMReX_Array4.H:283
__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:777
static constexpr bool IsArray4_v
True if this is an Array4 (N==4 and last dim is component).
Definition AMReX_Array4.H:290
__host__ __device__ U * ptr(Dim3 const &cell) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1163
__host__ __device__ constexpr bool ok() const noexcept
Check if the ArrayND pointer is valid and bounds are valid.
Definition AMReX_Array4.H:470
__host__ __device__ ArrayND(T *a_p, BoxND< N > const &box) noexcept
Constructor using a BoxND.
Definition AMReX_Array4.H:328
__host__ __device__ constexpr std::size_t size() const noexcept
Total number of elements in the ArrayND's index region.
Definition AMReX_Array4.H:638
__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:1190
T *__restrict__ p
Definition AMReX_Array4.H:292
__host__ __device__ constexpr bool contains(idx... i) const noexcept
Test whether an index tuple lies inside the ArrayND bounds.
Definition AMReX_Array4.H:684
__host__ __device__ T * ptr(idx... i) const noexcept
Multi-index ptr() for accessing pointer to element.
Definition AMReX_Array4.H:556
__host__ __device__ bool contains(IntVectND< M > const &iv) const noexcept
Test whether the spatial indices are inside the Array4 bounds.
Definition AMReX_Array4.H:1204
static constexpr bool IsLastDimComponent_v
True if the last dimension is treated as components.
Definition AMReX_Array4.H:288
__host__ __device__ CellData< T > cellData(int i, int j, int k) const noexcept
Create a single-cell component accessor.
Definition AMReX_Array4.H:762
__host__ __device__ constexpr ArrayND(ArrayND< std::remove_const_t< T >, N, last_dim_component > const &rhs) noexcept
Construct a const accessor from a non-const accessor.
Definition AMReX_Array4.H:317
__host__ __device__ constexpr int nComp() const noexcept
Get number of components.
Definition AMReX_Array4.H:621
__host__ __device__ bool contains(Dim3 const &cell) const noexcept
Test whether the spatial indices are inside the Array4 bounds.
Definition AMReX_Array4.H:1227
__host__ __device__ constexpr T * dataPtr() const noexcept
Get raw data pointer.
Definition AMReX_Array4.H:612
IntVectND< N > end
Exclusive upper bounds.
Definition AMReX_Array4.H:297
IntVectND< N > begin
Inclusive lower bounds.
Definition AMReX_Array4.H:296
__host__ __device__ U * ptr(int i, int j, int k, int n) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1088
__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:375
__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:357
__host__ __device__ constexpr Long get_stride() const noexcept
Return the stride (in elements) for dimension d.
Definition AMReX_Array4.H:662
__host__ __device__ constexpr ArrayND() noexcept
Default-construct an empty accessor.
Definition AMReX_Array4.H:305
__host__ __device__ U & operator()(idx... i) const noexcept
Multi-index operator() for accessing elements.
Definition AMReX_Array4.H:493
__host__ __device__ U * ptr(int i, int j, int k) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1060
__host__ __device__ U * ptr(Dim3 const &cell, int n) const noexcept
Return the pointer to an element.
Definition AMReX_Array4.H:1176
Lightweight accessor for data associated with a single cell.
Definition AMReX_Array4.H:42
__host__ __device__ constexpr CellData(T *a_p, Long a_stride, int a_ncomp)
Construct a CellData.
Definition AMReX_Array4.H:59
__host__ __device__ constexpr CellData(CellData< std::remove_const_t< T > > const &rhs) noexcept
Construct a const CellData from a non-const CellData.
Definition AMReX_Array4.H:74
__host__ __device__ U & operator[](int n) const noexcept
Access the n-th component of the cell.
Definition AMReX_Array4.H:105
__host__ __device__ int nComp() const noexcept
Return the number of components.
Definition AMReX_Array4.H:90
Definition AMReX_Dim3.H:12
Definition AMReX_Array4.H:1370
Definition AMReX_Array4.H:1384
__host__ __device__ PolymorphicArray4(Array4< T > const &a)
Definition AMReX_Array4.H:1386