Defined in header <cuda/atomic>:
template <typename T, cuda::thread_scope Scope = cuda::thread_scope_system> class cuda::atomic;
The class template cuda::atomic is an extended form of cuda::std::atomic that takes an additional cuda::thread_scope argument, defaulted to cuda::std::thread_scope_system.
It has the same interface and semantics as cuda::std::atomic, with the following additional operations.
Concurrency RestrictionsAn object of type cuda::atomic or cuda::std::atomic shall not be accessed concurrently by CPU and GPU threads unless:
Note, for objects of scopes other than cuda::thread_scope_system this is a data-race, and therefore also prohibited regardless of memory characteristics.
Under CUDA Compute Capability 6 (Pascal), an object of type atomic may not be used:
with automatic storage duration, or
if
is_always_lock_free()isfalse.
Under CUDA Compute Capability prior to 6 (Pascal), objects of type cuda::atomic or cuda::std::atomic may not be used.
For each type T and cuda::thread_scope S, the value of cuda::atomic<T, S>::is_always_lock_free() is as follows:
Type T
cuda::atomic<T, S>::is_always_lock_free()
Any valid type
Any thread scope
sizeof(T) <= 8
#include <cuda/atomic>
__global__ void example_kernel() {
// This atomic is suitable for all threads in the system.
cuda::atomic<int, cuda::thread_scope_system> a;
// This atomic has the same type as the previous one (`a`).
cuda::atomic<int> b;
// This atomic is suitable for all threads on the current processor (e.g. GPU).
cuda::atomic<int, cuda::thread_scope_device> c;
// This atomic is suitable for threads in the same thread block.
cuda::atomic<int, cuda::thread_scope_block> d;
}
RetroSearch is an open source project built by @garambo | Open a GitHub Issue
Search and Browse the WWW like it's 1997 | Search results from DuckDuckGo
HTML:
3.2
| Encoding:
UTF-8
| Version:
0.7.3