Kiedy używać lotnych ze współdzieloną pamięcią CUDA
W jakich okolicznościach należy używaćvolatile
słowo kluczowe z pamięcią współdzieloną jądra CUDA? Rozumiem, żevolatile
mówi kompilatorowi, aby nigdy nie buforował żadnych wartości, ale moje pytanie dotyczy zachowania ze wspólną tablicą:
__shared__ float products[THREADS_PER_ACTION];
// some computation
products[threadIdx.x] = localSum;
// wait for everyone to finish their computation
__syncthreads();
// then a (basic, ugly) reduction:
if (threadIdx.x == 0) {
float globalSum = 0.0f;
for (i = 0; i < THREADS_PER_ACTION; i++)
globalSum += products[i];
}
Czy potrzebujęproducts
być niestabilnym w tym przypadku? Dostęp do każdej pozycji tablicy jest możliwy tylko przez jeden wątek, z wyjątkiem końca, gdzie wszystko jest odczytywane przez wątek 0. Czy jest możliwe, że kompilator może buforować całą tablicę, a więc potrzebuję, aby byłvolatile
, czy będzie tylko buforować elementy?
Dzięki!