Atomics

Type-safe wrappers around the atomic operators found in GPU runtime APIs. More...

Functions
__device__ int	sp::atomicAdd (__device int *addr, int val)
	Atomically add val to the value stored at AS memory location addr, returning the original value. More...
__device__ unsigned int	sp::atomicAdd (__device unsigned int *addr, unsigned int val)
	Atomically add val to the value stored at AS memory location addr, returning the original value. More...
__device__ unsigned long long	sp::atomicAdd (__device unsigned long long *addr, unsigned long long val)
	Atomically add val to the value stored at AS memory location addr, returning the original value. More...
__device__ float	sp::atomicAdd (__device float *addr, float val)
	Atomically add val to the value stored at AS memory location addr, returning the original value. More...
__device__ double	sp::atomicAdd (__device double *addr, double val)
	Atomically add val to the value stored at AS memory location addr, returning the original value. More...
__device__ int	sp::atomicSub (__device int *addr, int val)
	Atomically subtract val from the value stored at AS memory location addr, returning the original value. More...
__device__ unsigned int	sp::atomicSub (__device unsigned int *addr, unsigned int val)
	Atomically subtract val from the value stored at AS memory location addr, returning the original value. More...
__device__ unsigned long long	sp::atomicSub (__device unsigned long long *addr, unsigned long long val)
	Atomically subtract val from the value stored at AS memory location addr, returning the original value. More...
__device__ float	sp::atomicSub (__device float *addr, float val)
	Atomically subtract val from the value stored at AS memory location addr, returning the original value. More...
__device__ double	sp::atomicSub (__device double *addr, double val)
	Atomically subtract val from the value stored at AS memory location addr, returning the original value. More...
__device__ int	sp::atomicExch (__device int *addr, int val)
	Atomically write val to addr and return the value that was stored there before calling this function. More...
__device__ unsigned int	sp::atomicExch (__device unsigned int *addr, unsigned int val)
	Atomically write val to addr and return the value that was stored there before calling this function. More...
__device__ unsigned long long	sp::atomicExch (__device unsigned long long *addr, unsigned long long val)
	Atomically write val to addr and return the value that was stored there before calling this function. More...
__device__ float	sp::atomicExch (__device float *addr, float val)
	Atomically write val to addr and return the value that was stored there before calling this function. More...
__device__ int	sp::atomicMin (__device int *addr, int val)
	Atomically write the min of *addr and val to addr, returning the original value of *addr. More...
__device__ int	sp::atomicMax (__device int *addr, int val)
	Atomically write the max of *addr and val to addr, returning the original value of *addr. More...
__device__ unsigned int	sp::atomicMin (__device unsigned int *addr, unsigned int val)
	Atomically write the min of *addr and val to addr, returning the original value of *addr. More...
__device__ unsigned int	sp::atomicMax (__device unsigned int *addr, unsigned int val)
	Atomically write the max of *addr and val to addr, returning the original value of *addr. More...
__device__ unsigned long long	sp::atomicMin (__device unsigned long long *addr, unsigned long long val)
	Atomically write the min of *addr and val to addr, returning the original value of *addr. More...
__device__ unsigned long long	sp::atomicMax (__device unsigned long long *addr, unsigned long long val)
	Atomically write the max of *addr and val to addr, returning the original value of *addr. More...
__device__ int	sp::atomicCAS (__device int *addr, int cmp, int val)
__device__ unsigned int	sp::atomicCAS (__device unsigned int *addr, unsigned int cmp, unsigned int val)
__device__ unsigned long long	sp::atomicCAS (__device unsigned long long *addr, unsigned long long cmp, unsigned long long val)
__device__ int	sp::atomicAnd (__device int *addr, int val)
	Atomically compute *addr = *addr & val and return the original value of *addr. More...
__device__ int	sp::atomicOr (__device int *addr, int val)
	Atomically compute *addr = *addr | val and return the original value of *addr. More...
__device__ int	sp::atomicXor (__device int *addr, int val)
	Atomically compute *addr = *addr ^ val and return the original value of *addr. More...
__device__ unsigned int	sp::atomicAnd (__device unsigned int *addr, unsigned int val)
	Atomically compute *addr = *addr & val and return the original value of *addr. More...
__device__ unsigned int	sp::atomicOr (__device unsigned int *addr, unsigned int val)
	Atomically compute *addr = *addr | val and return the original value of *addr. More...
__device__ unsigned int	sp::atomicXor (__device unsigned int *addr, unsigned int val)
	Atomically compute *addr = *addr ^ val and return the original value of *addr. More...
__device__ unsigned long long	sp::atomicAnd (__device unsigned long long *addr, unsigned long long val)
	Atomically compute *addr = *addr & val and return the original value of *addr. More...
__device__ unsigned long long	sp::atomicOr (__device unsigned long long *addr, unsigned long long val)
	Atomically compute *addr = *addr | val and return the original value of *addr. More...
__device__ unsigned long long	sp::atomicXor (__device unsigned long long *addr, unsigned long long val)
	Atomically compute *addr = *addr ^ val and return the original value of *addr. More...
__device__ unsigned int	sp::atomicInc (__device unsigned int *addr, unsigned int val)
__device__ unsigned int	sp::atomicDec (__device unsigned int *addr, unsigned int val)

Detailed Description

Type-safe wrappers around the atomic operators found in GPU runtime APIs.

The purpose of these wrappers is to add support for speclib address space qualifiers such as __device or __shared. This gives you additional type-safety when using atomics, since it is now impossible to accidentally pass a stack address, or incur a huge performance overhead by causing a generic atomic instruction to be used ( on many targets, these compile to conditional logic to select between address-space-specific atomic hardware instructions).

Function Documentation

atomicAdd() [1/5]

__device__ double sp::atomicAdd	(	__device double *	addr,
		double	val
	)

Atomically add val to the value stored at AS memory location addr, returning the original value.

atomicAdd() [2/5]

__device__ float sp::atomicAdd	(	__device float *	addr,
		float	val
	)

Atomically add val to the value stored at AS memory location addr, returning the original value.

atomicAdd() [3/5]

__device__ int sp::atomicAdd	(	__device int *	addr,
		int	val
	)

Atomically add val to the value stored at AS memory location addr, returning the original value.

atomicAdd() [4/5]

__device__ unsigned int sp::atomicAdd	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically add val to the value stored at AS memory location addr, returning the original value.

atomicAdd() [5/5]

__device__ unsigned long long sp::atomicAdd	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically add val to the value stored at AS memory location addr, returning the original value.

atomicAnd() [1/3]

__device__ int sp::atomicAnd	(	__device int *	addr,
		int	val
	)

Atomically compute *addr = *addr & val and return the original value of *addr.

atomicAnd() [2/3]

__device__ unsigned int sp::atomicAnd	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically compute *addr = *addr & val and return the original value of *addr.

atomicAnd() [3/3]

__device__ unsigned long long sp::atomicAnd	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically compute *addr = *addr & val and return the original value of *addr.

atomicExch() [1/4]

__device__ float sp::atomicExch	(	__device float *	addr,
		float	val
	)

Atomically write val to addr and return the value that was stored there before calling this function.

atomicExch() [2/4]

__device__ int sp::atomicExch	(	__device int *	addr,
		int	val
	)

Atomically write val to addr and return the value that was stored there before calling this function.

atomicExch() [3/4]

__device__ unsigned int sp::atomicExch	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically write val to addr and return the value that was stored there before calling this function.

atomicExch() [4/4]

__device__ unsigned long long sp::atomicExch	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically write val to addr and return the value that was stored there before calling this function.

atomicMax() [1/3]

__device__ int sp::atomicMax	(	__device int *	addr,
		int	val
	)

Atomically write the max of *addr and val to addr, returning the original value of *addr.

atomicMax() [2/3]

__device__ unsigned int sp::atomicMax	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically write the max of *addr and val to addr, returning the original value of *addr.

atomicMax() [3/3]

__device__ unsigned long long sp::atomicMax	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically write the max of *addr and val to addr, returning the original value of *addr.

atomicMin() [1/3]

__device__ int sp::atomicMin	(	__device int *	addr,
		int	val
	)

Atomically write the min of *addr and val to addr, returning the original value of *addr.

atomicMin() [2/3]

__device__ unsigned int sp::atomicMin	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically write the min of *addr and val to addr, returning the original value of *addr.

atomicMin() [3/3]

__device__ unsigned long long sp::atomicMin	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically write the min of *addr and val to addr, returning the original value of *addr.

atomicOr() [1/3]

__device__ int sp::atomicOr	(	__device int *	addr,
		int	val
	)

Atomically compute *addr = *addr | val and return the original value of *addr.

atomicOr() [2/3]

__device__ unsigned int sp::atomicOr	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically compute *addr = *addr | val and return the original value of *addr.

atomicOr() [3/3]

__device__ unsigned long long sp::atomicOr	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically compute *addr = *addr | val and return the original value of *addr.

atomicSub() [1/5]

__device__ double sp::atomicSub	(	__device double *	addr,
		double	val
	)

Atomically subtract val from the value stored at AS memory location addr, returning the original value.

atomicSub() [2/5]

__device__ float sp::atomicSub	(	__device float *	addr,
		float	val
	)

Atomically subtract val from the value stored at AS memory location addr, returning the original value.

atomicSub() [3/5]

__device__ int sp::atomicSub	(	__device int *	addr,
		int	val
	)

Atomically subtract val from the value stored at AS memory location addr, returning the original value.

atomicSub() [4/5]

__device__ unsigned int sp::atomicSub	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically subtract val from the value stored at AS memory location addr, returning the original value.

atomicSub() [5/5]

__device__ unsigned long long sp::atomicSub	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically subtract val from the value stored at AS memory location addr, returning the original value.

atomicXor() [1/3]

__device__ int sp::atomicXor	(	__device int *	addr,
		int	val
	)

Atomically compute *addr = *addr ^ val and return the original value of *addr.

atomicXor() [2/3]

__device__ unsigned int sp::atomicXor	(	__device unsigned int *	addr,
		unsigned int	val
	)

Atomically compute *addr = *addr ^ val and return the original value of *addr.

atomicXor() [3/3]

__device__ unsigned long long sp::atomicXor	(	__device unsigned long long *	addr,
		unsigned long long	val
	)

Atomically compute *addr = *addr ^ val and return the original value of *addr.