CUDA CURAND是否容易受數據競爭影響？

Question

我產卵的從我Setup()核心256個線程設置了256個CURAND狀態的陣列RNGstates1塊：

__global__ void Setup(curandState *RNGstates, long seed) { 
    int tid = threadIdx.x; 
    curand_init(seed, tid, 0, &RNGstates[tid]); 
} 

現在，我產卵從我Generate()內核的1000塊256個線程填寫數組result與256,000個隨機數。但是，我這樣做只使用RNGstates的256個狀態，使得每個國家將通過1000線（每個塊）訪問：

__global__ void Generate(curandState *RNGstates, float *result) { 
    int tid = blockIdx.x*blockDim.x + threadIdx.x; 
    float rnd = curand_uniform(&RNGstates[threadIdx.x]); 
    result[tid] = rnd; 
} 

我知道打電話curand_uniform()莫名其妙地更新狀態，所以我相信一些寫操作正在發生。

因此，當1000個線程映射到256個CURAND狀態中的每個狀態時，都應該擔心數據競爭會通過curand_uniform()隱式更新狀態？這會影響我的隨機數的質量（例如獲得頻繁的重複值）嗎？

非常感謝。

Answer 1

我認爲分享國家肯定會影響質量。重複值是共享狀態的最佳狀態。數據競賽可能會徹底毀滅各州。

您可以爲每個線程保留一個狀態。

使用1000個塊時，您的情況需要256,000個狀態。該代碼應該像

__global__ void Setup(curandState *RNGstates, long seed) { 
    int tid = blockIdx.x*blockDim.x + threadIdx.x; 
    curand_init(seed, tid, 0, &RNGstates[tid]); 
} 

和

__global__ void Generate(curandState *RNGstates, float *result) { 
    int tid = blockIdx.x*blockDim.x + threadIdx.x; 
    float rnd = curand_uniform(&RNGstates[tid]); 
    result[tid] = rnd; 
} 

爲了減少多塊MEM的要求，你可以在你的＃封鎖限制在一個小數目，併產生每個線程多個隨機數，而不是1每個線程的隨機數。

__global__ void generate_uniform_kernel(curandState *state, 
           unsigned int *result) 
{ 
    int id = threadIdx.x + blockIdx.x * 64; 
    unsigned int count = 0; 
    float x; 
    /* Copy state to local memory for efficiency */ 
    curandState localState = state[id]; 
    /* Generate pseudo-random uniforms */ 
    for(int n = 0; n < 10000; n++) { 
     x = curand_uniform(&localState); 
     /* Check if > .5 */ 
     if(x > .5) { 
      count++; 
     } 
    } 
    /* Copy state back to global memory */ 
    state[id] = localState; 
    /* Store results */ 
    result[id] += count; 
} 

請參閱有關如何處理多發塊完整的例子cuRAND參考手冊中的部分Device API Examples。

Answer 2

你也可以在你需要每塊只有一個（如果塊在每個大多數256個線程）http://docs.nvidia.com/cuda/curand/device-api-overview.html#bit-generation-1