Block-Structured AMR Software Framework
Loading...
Searching...
No Matches
AMReX_GpuLaunchMacrosC.nolint.H File Reference

Go to the source code of this file.

Macros

#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE(TN, TI, block)
#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_2(TN1, TI1, block1, TN2, TI2, block2)
#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_3(TN1, TI1, block1, TN2, TI2, block2, TN3, TI3, block3)
#define AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE(...)
#define AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE_2(...)
#define AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE_3(...)
#define AMREX_FOR_1D_IMPL(n, i, block)
#define AMREX_PARALLEL_FOR_1D_IMPL(n, i, block)
#define AMREX_FOR_3D_IMPL(box, i, j, k, block)
#define AMREX_PARALLEL_FOR_3D_IMPL(box, i, j, k, block)
#define AMREX_FOR_4D_IMPL(box, ncomp, i, j, k, n, block)
#define AMREX_PARALLEL_FOR_4D_IMPL(box, ncomp, i, j, k, n, block)
#define AMREX_FOR_1D(...)
#define AMREX_FOR_3D(...)
#define AMREX_FOR_4D(...)
#define AMREX_PARALLEL_FOR_1D(...)
#define AMREX_PARALLEL_FOR_3D(...)
#define AMREX_PARALLEL_FOR_4D(...)
#define AMREX_HOST_DEVICE_FOR_1D(...)
#define AMREX_HOST_DEVICE_FOR_3D(...)
#define AMREX_HOST_DEVICE_FOR_4D(...)
#define AMREX_HOST_DEVICE_PARALLEL_FOR_1D(...)
#define AMREX_HOST_DEVICE_PARALLEL_FOR_3D(...)
#define AMREX_HOST_DEVICE_PARALLEL_FOR_4D(...)

Macro Definition Documentation

◆ AMREX_FOR_1D

#define AMREX_FOR_1D ( ...)
Value:
AMREX_FOR_1D_IMPL(__VA_ARGS__);
AMREX_FOR_1D_IMPL
#define AMREX_FOR_1D_IMPL(n, i, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:38

◆ AMREX_FOR_1D_IMPL

#define AMREX_FOR_1D_IMPL ( n,
i,
block )
Value:
for (auto i = decltype(n){0}; i < n; ++i) { \
block \
}

◆ AMREX_FOR_3D

#define AMREX_FOR_3D ( ...)
Value:
AMREX_FOR_3D_IMPL(__VA_ARGS__);
AMREX_FOR_3D_IMPL
#define AMREX_FOR_3D_IMPL(box, i, j, k, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:49

◆ AMREX_FOR_3D_IMPL

#define AMREX_FOR_3D_IMPL ( box,
i,
j,
k,
block )
Value:
{ \
const auto amrex_i_lo = amrex::lbound(box); \
const auto amrex_i_hi = amrex::ubound(box); \
for (int k = amrex_i_lo.z; k <= amrex_i_hi.z; ++k) { \
for (int j = amrex_i_lo.y; j <= amrex_i_hi.y; ++j) { \
for (int i = amrex_i_lo.x; i <= amrex_i_hi.x; ++i) { \
block \
}}} \
}
amrex::ubound
__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
amrex::lbound
__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

◆ AMREX_FOR_4D

#define AMREX_FOR_4D ( ...)
Value:
AMREX_FOR_4D_IMPL(__VA_ARGS__);
AMREX_FOR_4D_IMPL
#define AMREX_FOR_4D_IMPL(box, ncomp, i, j, k, n, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:72

◆ AMREX_FOR_4D_IMPL

#define AMREX_FOR_4D_IMPL ( box,
ncomp,
i,
j,
k,
n,
block )
Value:
{ \
const auto amrex_i_lo = amrex::lbound(box); \
const auto amrex_i_hi = amrex::ubound(box); \
for (int n = 0; n < ncomp; ++n) { \
for (int k = amrex_i_lo.z; k <= amrex_i_hi.z; ++k) { \
for (int j = amrex_i_lo.y; j <= amrex_i_hi.y; ++j) { \
for (int i = amrex_i_lo.x; i <= amrex_i_hi.x; ++i) { \
block \
}}}} \
}

◆ AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE

#define AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE ( ...)
Value:
AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE(__VA_ARGS__)
AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE
#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE(TN, TI, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:4

◆ AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE_2

#define AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE_2 ( ...)
Value:
AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_2(__VA_ARGS__)
AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_2
#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_2(TN1, TI1, block1, TN2, TI2, block2)
Definition AMReX_GpuLaunchMacrosC.nolint.H:11

◆ AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE_3

#define AMREX_GPU_LAUNCH_DEVICE_LAMBDA_RANGE_3 ( ...)
Value:
AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_3(__VA_ARGS__)
AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_3
#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_3(TN1, TI1, block1, TN2, TI2, block2, TN3, TI3, block3)
Definition AMReX_GpuLaunchMacrosC.nolint.H:21

◆ AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE

#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE ( TN,
TI,
block )
Value:
{ \
for (auto const TI : amrex::Gpu::Range(TN)) { \
block \
} \
}
amrex::Gpu::Range
__host__ __device__ range_detail::range_impl< T > Range(T const &b) noexcept
Definition AMReX_GpuRange.H:128

◆ AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_2

#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_2 ( TN1,
TI1,
block1,
TN2,
TI2,
block2 )
Value:
{ \
for (auto const TI1 : amrex::Gpu::Range(TN1)) { \
block1 \
} \
for (auto const TI2 : amrex::Gpu::Range(TN2)) { \
block2 \
} \
}

◆ AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_3

#define AMREX_GPU_LAUNCH_HOST_DEVICE_LAMBDA_RANGE_3 ( TN1,
TI1,
block1,
TN2,
TI2,
block2,
TN3,
TI3,
block3 )
Value:
{ \
for (auto const TI1 : amrex::Gpu::Range(TN1)) { \
block1 \
} \
for (auto const TI2 : amrex::Gpu::Range(TN2)) { \
block2 \
} \
for (auto const TI3 : amrex::Gpu::Range(TN3)) { \
block3 \
} \
}

◆ AMREX_HOST_DEVICE_FOR_1D

#define AMREX_HOST_DEVICE_FOR_1D ( ...)
Value:
AMREX_FOR_1D_IMPL(__VA_ARGS__);

◆ AMREX_HOST_DEVICE_FOR_3D

#define AMREX_HOST_DEVICE_FOR_3D ( ...)
Value:
AMREX_FOR_3D_IMPL(__VA_ARGS__);

◆ AMREX_HOST_DEVICE_FOR_4D

#define AMREX_HOST_DEVICE_FOR_4D ( ...)
Value:
AMREX_FOR_4D_IMPL(__VA_ARGS__);

◆ AMREX_HOST_DEVICE_PARALLEL_FOR_1D

#define AMREX_HOST_DEVICE_PARALLEL_FOR_1D ( ...)
Value:
AMREX_PARALLEL_FOR_1D_IMPL(__VA_ARGS__);
AMREX_PARALLEL_FOR_1D_IMPL
#define AMREX_PARALLEL_FOR_1D_IMPL(n, i, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:43

◆ AMREX_HOST_DEVICE_PARALLEL_FOR_3D

#define AMREX_HOST_DEVICE_PARALLEL_FOR_3D ( ...)
Value:
AMREX_PARALLEL_FOR_3D_IMPL(__VA_ARGS__);
AMREX_PARALLEL_FOR_3D_IMPL
#define AMREX_PARALLEL_FOR_3D_IMPL(box, i, j, k, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:60

◆ AMREX_HOST_DEVICE_PARALLEL_FOR_4D

#define AMREX_HOST_DEVICE_PARALLEL_FOR_4D ( ...)
Value:
AMREX_PARALLEL_FOR_4D_IMPL(__VA_ARGS__);
AMREX_PARALLEL_FOR_4D_IMPL
#define AMREX_PARALLEL_FOR_4D_IMPL(box, ncomp, i, j, k, n, block)
Definition AMReX_GpuLaunchMacrosC.nolint.H:84

◆ AMREX_PARALLEL_FOR_1D

#define AMREX_PARALLEL_FOR_1D ( ...)
Value:
AMREX_PARALLEL_FOR_1D_IMPL(__VA_ARGS__);

◆ AMREX_PARALLEL_FOR_1D_IMPL

#define AMREX_PARALLEL_FOR_1D_IMPL ( n,
i,
block )
Value:
AMREX_PRAGMA_SIMD \
for (auto i = decltype(n){0}; i < n; ++i) { \
block \
}

◆ AMREX_PARALLEL_FOR_3D

#define AMREX_PARALLEL_FOR_3D ( ...)
Value:
AMREX_PARALLEL_FOR_3D_IMPL(__VA_ARGS__);

◆ AMREX_PARALLEL_FOR_3D_IMPL

#define AMREX_PARALLEL_FOR_3D_IMPL ( box,
i,
j,
k,
block )
Value:
{ \
const auto amrex_i_lo = amrex::lbound(box); \
const auto amrex_i_hi = amrex::ubound(box); \
for (int k = amrex_i_lo.z; k <= amrex_i_hi.z; ++k) { \
for (int j = amrex_i_lo.y; j <= amrex_i_hi.y; ++j) { \
AMREX_PRAGMA_SIMD \
for (int i = amrex_i_lo.x; i <= amrex_i_hi.x; ++i) { \
block \
}}} \
}

◆ AMREX_PARALLEL_FOR_4D

#define AMREX_PARALLEL_FOR_4D ( ...)
Value:
AMREX_PARALLEL_FOR_4D_IMPL(__VA_ARGS__);

◆ AMREX_PARALLEL_FOR_4D_IMPL

#define AMREX_PARALLEL_FOR_4D_IMPL ( box,
ncomp,
i,
j,
k,
n,
block )
Value:
{ \
const auto amrex_i_lo = amrex::lbound(box); \
const auto amrex_i_hi = amrex::ubound(box); \
for (int n = 0; n < ncomp; ++n) { \
for (int k = amrex_i_lo.z; k <= amrex_i_hi.z; ++k) { \
for (int j = amrex_i_lo.y; j <= amrex_i_hi.y; ++j) { \
AMREX_PRAGMA_SIMD \
for (int i = amrex_i_lo.x; i <= amrex_i_hi.x; ++i) { \
block \
}}}} \
}