atomicAdd with signed long long not working

Accelerated Computing CUDA CUDA Programming and Performance
1

Hi,

I am trying to add an integer to a (signed) long long global variable using the atomicAdd function using Cuda 3.2. Unfortunately atomicAdd takes unsigned long long types and no signed long long types. The Cuda C Programming Guide says that any atomic operation can be implemented based on atomicCAS(). Which is what I am trying, but I am not succesfull. Below is my code:

__device__ long long atomicAdd(long long* address, int val) 

{

	unsigned long long int* address_as_ull = (unsigned long long int*)address;

	unsigned long long int old = *address_as_ull, assumed;

	

	do 

	{ 

		assumed = old;

		old = atomicCAS(address_as_ull, assumed, val + assumed);

	} while (assumed != old); 

	

	return old;

}

I am also unable to print long long variables to my console. I have tried cuPrintf(“%lld”, long_value), cuPrintf(“%d”, long_value) and cuPrintf(“%1f”, long_value) but none of them are working.

Can anyone suggest a function that does work or another way to add an integer to a signed long long variable in global memory in an atomic fashion. Thank you very much in advance.

Kind regards,

Nick

2

There is no need to use atomicCAS(). Just use atomicAdd(), casting its arguments to (unsigned long long*), (unsigned long long).

Signed and unsigned add give bit-identical results. The difference is only how the result is interpreted.

3

Thank you very much for the help. The casting worked perfectly. I’m glad it was this easy. Thanks!