Est-il sûr d'utiliser __syncthreads()
dans un bloc où j'ai volontairement supprimé des threads en utilisant return
?
La documentation indique que __syncthreads()
doit être appelé par chaque thread du bloc, sinon cela conduira à un blocage, mais en pratique, je n'ai jamais rencontré un tel comportement.
Exemple de code :
__global__ void kernel(float* data, size_t size) {
// Drop excess threads if user put too many in kernel call.
// After the return, there are `size` active threads.
if (threadIdx.x >= size) {
return;
}
// ... do some work ...
__syncthreads(); // Is this safe?
// For the rest of the kernel, we need to drop one excess thread
// After the return, there are `size - 1` active threads
if (threadIdx.x + 1 == size) {
return;
}
// ... do more work ...
__syncthreads(); // Is this safe?
}