小编use*_*789的帖子

在运行以下内核时,我有以下容易重现的问题,除了atomicAdds of floats之外什么都不做:

#define OUT_ITERATIONS 20000000
#define BLOCKS 12
#define THREADS 192

__global__ void testKernel(float* result) {
    int i = threadIdx.x + blockIdx.x * blockDim.x;
    float bias = 1.0f;
    int n = 1;

    while (i < OUT_ITERATIONS) {
        atomicAdd(result, bias);
        i += BLOCKS * THREADS;
    }
}

内核应该将结果递增OUT_ITERATIONS次数,即20M.我用这个标准代码调用内核:

int main() {
cudaError_t cudaStatus;
float* result;
float* dev_result;

// Choose which GPU to run on, change this on a multi-GPU system.
cudaStatus = cudaSetDevice(0);
if (cudaStatus != cudaSuccess) {
    fprintf(stderr, "cudaSetDevice …

当我float atomicAdd(float *address, float val)用来添加小于约的浮点值时.1e-39to 0,添加不起作用,并且值address仍为0.

这是最简单的代码:

__device__ float test[6] = {0};
__global__ void testKernel() {
    float addit = sinf(1e-20);
    atomicAdd(&test[0], addit);
    test[1] += addit;
    addit = sinf(1e-37);
    atomicAdd(&test[2], addit);
    test[3] += addit;
    addit = sinf(1e-40);
    atomicAdd(&test[4], addit);
    test[5] += addit;
}

当我运行上面的代码testKernel<<<1, 1>>>();并停止使用调试器时,我看到:

test    0x42697800
    [0] 9.9999997e-21
    [1] 9.9999997e-21
    [2] 9.9999999e-38
    [3] 9.9999999e-38
    [4] 0            
    [5] 9.9999461e-41

注意test [4]和test [5]之间的区别.两者都做了同样的事情,但简单的添加工作,原子一个什么都没做.我在这里错过了什么？

更新:系统信息:CUDA 5.5.20,NVidia Titan卡,驱动程序331.82,Windows 7x64,Nsight 3.2.1.13309.