Block-Structured AMR Software Framework
Loading...
Searching...
No Matches
amrex::Gpu::Atomic Namespace Reference

Functions

template<class T>
__device__ AMREX_FORCE_INLINE T Add_device (T *const sum, T const value) noexcept
__device__ AMREX_FORCE_INLINE Long Add_device (Long *const sum, Long const value) noexcept
template<class T>
__host__ __device__ AMREX_FORCE_INLINE T Add (T *sum, T value) noexcept
template<typename T, typename Op, typename Cond>
requires (sizeof(T) == sizeof(unsigned int))
__device__ AMREX_FORCE_INLINE bool If_device (T *const sum, T const value, Op &&op, Cond &&cond) noexcept
template<typename T, typename Op, typename Cond>
requires (sizeof(T) == sizeof(unsigned long long))
__device__ AMREX_FORCE_INLINE bool If_device (T *const sum, T const value, Op &&op, Cond &&cond) noexcept
template<class T, class Op, class Cond>
__host__ __device__ AMREX_FORCE_INLINE bool If (T *const add, T const value, Op &&op, Cond &&cond) noexcept
 Conditionally perform an atomic operation.
template<class T>
__host__ __device__ AMREX_FORCE_INLINE void AddNoRet (T *sum, T value) noexcept
template<class T>
__device__ AMREX_FORCE_INLINE T Min_device (T *const m, T const value) noexcept
__device__ AMREX_FORCE_INLINE float Min_device (float *const m, float const value) noexcept
__device__ AMREX_FORCE_INLINE double Min_device (double *const m, double const value) noexcept
__device__ AMREX_FORCE_INLINE Long Min_device (Long *const m, Long const value) noexcept
template<class T>
__host__ __device__ AMREX_FORCE_INLINE T Min (T *const m, T const value) noexcept
template<class T>
__device__ AMREX_FORCE_INLINE T Max_device (T *const m, T const value) noexcept
__device__ AMREX_FORCE_INLINE float Max_device (float *const m, float const value) noexcept
__device__ AMREX_FORCE_INLINE double Max_device (double *const m, double const value) noexcept
__device__ AMREX_FORCE_INLINE Long Max_device (Long *const m, Long const value) noexcept
template<class T>
__host__ __device__ AMREX_FORCE_INLINE T Max (T *const m, T const value) noexcept
__host__ __device__ AMREX_FORCE_INLINE int LogicalOr (int *const m, int const value) noexcept
__host__ __device__ AMREX_FORCE_INLINE int LogicalAnd (int *const m, int const value) noexcept
template<typename T>
__host__ __device__ AMREX_FORCE_INLINE T Exch (T *address, T val) noexcept
template<typename T>
__host__ __device__ AMREX_FORCE_INLINE T CAS (T *const address, T compare, T const val) noexcept
template<typename T>
requires (sizeof(T) == sizeof(int))
__device__ AMREX_FORCE_INLINE T Multiply_device (T *const prod, T const value) noexcept
template<typename T>
requires (sizeof(T) == sizeof(unsigned long long))
__device__ AMREX_FORCE_INLINE T Multiply_device (T *const prod, T const value) noexcept
template<class T>
__host__ __device__ AMREX_FORCE_INLINE T Multiply (T *const prod, T const value) noexcept
template<typename T>
requires (sizeof(T) == sizeof(int))
__device__ AMREX_FORCE_INLINE T Divide_device (T *const quot, T const value) noexcept
template<typename T>
requires (sizeof(T) == sizeof(unsigned long long))
__device__ AMREX_FORCE_INLINE T Divide_device (T *const quot, T const value) noexcept
template<class T>
__host__ __device__ AMREX_FORCE_INLINE T Divide (T *const quot, T const value) noexcept

Function Documentation

◆ Add()

template<class T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Add ( T * sum,
T value )
noexcept

◆ Add_device() [1/2]

__device__ AMREX_FORCE_INLINE Long amrex::Gpu::Atomic::Add_device ( Long *const sum,
Long const value )
noexcept

◆ Add_device() [2/2]

template<class T>
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Add_device ( T *const sum,
T const value )
noexcept

◆ AddNoRet()

template<class T>
__host__ __device__ AMREX_FORCE_INLINE void amrex::Gpu::Atomic::AddNoRet ( T * sum,
T value )
noexcept

◆ CAS()

template<typename T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::CAS ( T *const address,
T compare,
T const val )
noexcept

◆ Divide()

template<class T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Divide ( T *const quot,
T const value )
noexcept

◆ Divide_device() [1/2]

template<typename T>
requires (sizeof(T) == sizeof(unsigned long long))
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Divide_device ( T *const quot,
T const value )
noexcept

◆ Divide_device() [2/2]

template<typename T>
requires (sizeof(T) == sizeof(int))
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Divide_device ( T *const quot,
T const value )
noexcept

◆ Exch()

template<typename T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Exch ( T * address,
T val )
noexcept

◆ If()

template<class T, class Op, class Cond>
__host__ __device__ AMREX_FORCE_INLINE bool amrex::Gpu::Atomic::If ( T *const add,
T const value,
Op && op,
Cond && cond )
noexcept

Conditionally perform an atomic operation.

Atomically updates the result at "add" with "value" using "op", but only if "cond" is true.

Template Parameters
Tthe type pointed to by add
Opcallable that takes two T argument and combines them
Condcallable that takes a "T" a returns whether to do the update
Parameters
addaddress to atomically update
valuevalue to combine
opcallable specifying the operation to use to combine *add and value
condcallable specifying the condition to test on first. The value passed in to the cond function is the would-be combined value

◆ If_device() [1/2]

template<typename T, typename Op, typename Cond>
requires (sizeof(T) == sizeof(unsigned long long))
__device__ AMREX_FORCE_INLINE bool amrex::Gpu::Atomic::If_device ( T *const sum,
T const value,
Op && op,
Cond && cond )
noexcept

◆ If_device() [2/2]

template<typename T, typename Op, typename Cond>
requires (sizeof(T) == sizeof(unsigned int))
__device__ AMREX_FORCE_INLINE bool amrex::Gpu::Atomic::If_device ( T *const sum,
T const value,
Op && op,
Cond && cond )
noexcept

◆ LogicalAnd()

__host__ __device__ AMREX_FORCE_INLINE int amrex::Gpu::Atomic::LogicalAnd ( int *const m,
int const value )
noexcept

◆ LogicalOr()

__host__ __device__ AMREX_FORCE_INLINE int amrex::Gpu::Atomic::LogicalOr ( int *const m,
int const value )
noexcept

◆ Max()

template<class T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Max ( T *const m,
T const value )
noexcept

◆ Max_device() [1/4]

__device__ AMREX_FORCE_INLINE double amrex::Gpu::Atomic::Max_device ( double *const m,
double const value )
noexcept

◆ Max_device() [2/4]

__device__ AMREX_FORCE_INLINE float amrex::Gpu::Atomic::Max_device ( float *const m,
float const value )
noexcept

◆ Max_device() [3/4]

__device__ AMREX_FORCE_INLINE Long amrex::Gpu::Atomic::Max_device ( Long *const m,
Long const value )
noexcept

◆ Max_device() [4/4]

template<class T>
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Max_device ( T *const m,
T const value )
noexcept

◆ Min()

template<class T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Min ( T *const m,
T const value )
noexcept

◆ Min_device() [1/4]

__device__ AMREX_FORCE_INLINE double amrex::Gpu::Atomic::Min_device ( double *const m,
double const value )
noexcept

◆ Min_device() [2/4]

__device__ AMREX_FORCE_INLINE float amrex::Gpu::Atomic::Min_device ( float *const m,
float const value )
noexcept

◆ Min_device() [3/4]

__device__ AMREX_FORCE_INLINE Long amrex::Gpu::Atomic::Min_device ( Long *const m,
Long const value )
noexcept

◆ Min_device() [4/4]

template<class T>
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Min_device ( T *const m,
T const value )
noexcept

◆ Multiply()

template<class T>
__host__ __device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Multiply ( T *const prod,
T const value )
noexcept

◆ Multiply_device() [1/2]

template<typename T>
requires (sizeof(T) == sizeof(unsigned long long))
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Multiply_device ( T *const prod,
T const value )
noexcept

◆ Multiply_device() [2/2]

template<typename T>
requires (sizeof(T) == sizeof(int))
__device__ AMREX_FORCE_INLINE T amrex::Gpu::Atomic::Multiply_device ( T *const prod,
T const value )
noexcept