c - CUDA 中的位数组

Question

我在 CUDA 中实现了埃拉托斯特尼筛法，并且得到了一个非常奇怪的输出。我使用 unsigned char* 作为数据结构，并使用以下宏来操作位。

#define ISBITSET(x,i) ((x[i>>3] & (1<<(i&7)))!=0)
#define SETBIT(x,i) x[i>>3]|=(1<<(i&7));
#define CLEARBIT(x,i) x[i>>3]&=(1<<(i&7))^0xFF;

我设置该位来表示它是素数，否则它 = 0。这是我调用内核的地方

size_t p=3;
size_t primeTill = 30;

while(p*p<=primeTill)
{
    if(ISBITSET(h_a, p) == 1){
        int dimA = 30;
        int numBlocks = 1;
        int numThreadsPerBlock = dimA;
        dim3 dimGrid(numBlocks);
        dim3 dimBlock(numThreadsPerBlock);
        cudaMemcpy( d_a, h_a, memSize, cudaMemcpyHostToDevice );        
        cudaThreadSynchronize();    
        reverseArrayBlock<<< dimGrid, dimBlock >>>( d_a, primeTill, p );
        cudaThreadSynchronize();    
        cudaMemcpy( h_a, d_a, memSize, cudaMemcpyDeviceToHost );
        cudaThreadSynchronize();    
        printf("This is after removing multiples of %d\n", p);
        //Loop
        for(size_t i = 0; i < primeTill +1; i++)
        {
            printf("Bit %d is %d\n", i, ISBITSET(h_a, i));
        }
    }           
    p++;
}

这是我的内核

__global__ void reverseArrayBlock(unsigned char *d_out, int size, size_t p)
{
int id = blockIdx.x*blockDim.x + threadIdx.x;
int r = id*p;
if(id >= p && r <= size )
{
    while(ISBITSET(d_out, r ) == 1 ){
        CLEARBIT(d_out, r);
    }

    // if(r == 9)
    // {
    //  /* code */
    //  CLEARBIT(d_out, 9);
    // }

}

}输出应该是:2, 3, 5, 7, 11, 13, 17, 19, 23, 29而我的输出是:2、3、5、9、7、11、13、17、19、23、29

如果你看一下内核代码，如果我取消注释这些行，我将得到正确的答案，这意味着我的循环或检查没有任何问题!

Answer 1

多个线程同时访问全局内存中的同一个字(字符)，因此写入的结果会被损坏。

您可以使用原子操作来防止这种情况，但更好的解决方案是改变您的算法:不要让每个线程筛选 2、3、4、5... 的倍数，而是让每个线程检查一个范围，如 [ 0..7]、[8..15]、...，这样每个范围的长度都是 8 位的倍数，并且不会发生冲突。