我使用的是具有 1.3 计算能力和 nvcc 编译器驱动程序 4.0 的 Tesla C1060。我正在尝试对线程 block 进行一些本地计算。每个线程 block 都提供了一个共享数组,该数组首先被初始化为零值。为了同步线程 block 的线程对共享数据的并发更新(添加),我使用了 CUDA atomicAdd 原语。
一旦每个线程 block 准备好其共享数据数组中的结果,共享数据数组中的每个条目将迭代合并(使用 atomicAdd)到全局数据数组中的相应条目。
下面是一段与我基本上想做的非常相似的代码。
#define DATA_SZ 16
typedef unsigned long long int ULLInt;
__global__ void kernel( ULLInt* data, ULLInt ThreadCount )
{
ULLInt thid = threadIdx.x + blockIdx.x * blockDim.x;
__shared__ ULLInt sharedData[DATA_SZ];
// Initialize the shared data
if( threadIdx.x == 0 )
{
for( int i = 0; i < DATA_SZ; i++ ) { sharedData[i] = 0; }
}
__syncthreads();
//..some code here
if( thid < ThreadCount )
{
//..some code here
atomicAdd( &sharedData[getIndex(thid), thid );
//..some code here
for(..a loop...)
{
//..some code here
if(thid % 2 == 0)
{
// getIndex() returns a value in [0, DATA_SZ )
atomicAdd( &sharedData[getIndex(thid)], thid * thid );
}
}
}
__syncthreads();
if( threadIdx.x == 0 )
{
// ...
for( int i = 0; i < DATA_SZ; i++ ) { atomicAdd( &Data[i], sharedData[i] ); }
//...
}
}
如果我使用 -arch=sm_20 编译,我不会收到任何错误。但是,当我使用 -arch=sm_13 选项编译内核时,出现以下错误:
ptxas /tmp/tmpxft_00004dcf_00000000-2_mycode.ptx, line error : Global state space expected for instruction 'atom'
ptxas /tmp/tmpxft_00004dcf_00000000-2_mycode.ptx, line error : Global state space expected for instruction 'atom'
ptxas fatal : Ptx assembly aborted due to errors
如果我评论以下两行,我不会收到任何关于 -arch=sm_13 的错误:
atomicAdd( &sharedData[getIndex(thid), thid );
atomicAdd( &sharedData[getIndex(thid)], thid * thid );
有人可以建议我可能做错了什么吗?
最佳答案
在CUDA C编程指南中找到解决方法:Atomic functions operation on shared memory and atomic functions operation on 64-bit words only available for devices of compute capability 1.2 and above.对共享内存中的 64 位字进行操作的原子函数仅适用于计算能力 2.x 及更高版本的设备。
所以基本上我不能在这里使用 ULLInt 共享内存,不知何故我需要使用 unsigned int
关于c++ - CUDA共享内存原子错误,我们在Stack Overflow上找到一个类似的问题: https://stackoverflow.com/questions/12150615/