無効なデバイスシンボルcudaMemcpyFromSymbol CUDA

Question

CUDAで配列のすべての要素の合計を計算します。私はこのコードを思いついた。それはエラーなしでコンパイルされます。しかし、結果は常にゼロです。無効なデバイスシンボルはcudaMemcpyFromSymbolです。スラストやクライブのようなライブラリは使えません。

#define TRIALS_PER_THREAD 4096 
#define NUM_BLOCKS 256 
#define NUM_THREADS 256 
double *dev; 
__device__ volatile double pi_gpu = 0; 

__global__ void ArraySum(double *array) 

{ 
unsigned int tid = threadIdx.x + blockDim.x * blockIdx.x; 
pi_gpu = pi_gpu + array[tid]; 
__syncthreads(); 
} 

int main (int argc, char *argv[]) { 
cudaMalloc((void **) &dev, NUM_BLOCKS * NUM_THREADS * sizeof(double)); 
    double pi_gpu_h; 

ArraySum<<<NUM_BLOCKS, NUM_THREADS>>>(dev); 
cudaDeviceSynchronize(); 
cudaError err = cudaMemcpyFromSymbol(&pi_gpu_h, &pi_gpu, sizeof(double), cudaMemcpyDeviceToHost); 
if(cudaSuccess != err) 
{ 
    fprintf(stderr, "cudaMemcpyFromSymbolfailed : %s\n", cudaGetErrorString(err)); 
    exit(-1); 
} 
return pi_gpu_h; // this is always zero!!! 
} 

Answer 1

コードはスレッドセーフではありません。複数のスレッドからグローバル変数に書き込むことはまったく安全ではありません。削減カーネルがことができるかのこの例では：

//Untested code 
global_void plus_reduce(int *input, int N, int *total){ 
    int tid = threadIdx.x; 
    int i = blockIdx.x*blockDim.x + threadIdx.x; 
    // Each block loads its elements into shared memory 
    _shared_ int x[blocksize]; 
    x[tid] = (i<N) ? input[i] : 0; // last block may pad with 0’s 
    _syncthreads(); 
    // Build summation tree over elements. 
    for(int s=blockDim.x/2; s>0; s=s/2){ 
     if(tid < s) x[tid] += x[tid + s]; 
    _syncthreads(); 
} 
// Thread 0 adds the partial sum to the total sum 
if(tid == 0) 
    atomicAdd(total, x[tid]);     
} 

Source

Answer 2

シンボルコールからのコピーでシンボル引数が正しくありません。

cudaMemcpyFromSymbol(&pi_gpu_h, pi_gpu, sizeof(double), cudaMemcpyDeviceToHost)