атомные операции в CUDA

Question

Следующая программа использовала реализацию атомарных блокировок из 'Cuda By Example', но при запуске программы мой компьютер завис. Может кто-нибудь сказать мне, что не так с моей программой? Большое спасибо

Ифэй

#include <stdio.h>


__global__ void test()
{
    __shared__ int i, mutex;

    if (threadIdx.x == 0) {
       i = 0;
       mutex = 0;
    }
    __syncthreads();

    while( atomicCAS(&mutex, 0, 1) != 0);
    i++;
    printf("thread %d: %d\n", threadIdx.x, i);
    atomicExch(&mutex,0);
}

Answer 1

Вот теория. Я надеюсь, что вы знакомы с концепцией деформации. В цикле while все потоки внутри деформации войдут в цикл while. Выходит только один, а остальные потоки будут находиться внутри цикла while. Это представит расходящуюся ветвь, делающую поток, который вышел из цикла while, останавливается, пока ветвь не сходится снова. Поскольку этот поток является единственным, который может освободить мьютекс, этого никогда не произойдет, поскольку он ожидает, пока другие потоки не сойдутся.