ungültiges Gerätesymbol cudaMemcpyFromSymbol CUDA

Question

Ich möchte die Summe aller Elemente eines Arrays in CUDA berechnen. Ich habe diesen Code entwickelt. Es kompiliert ohne Fehler. Aber das Ergebnis ist immer Null. Ich habe das ungültige Gerätesymbol von cudaMemcpyFromSymbol. Ich kann keine Bibliotheken wie Thrust oder Cublas verwenden.

#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

Ihr Code ist nicht threadsicher. Das Schreiben in eine globale Variable aus mehreren Threads ist nicht sicher. Dieses Beispiel dafür, wie Reduktion Kernel kann sein:

//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

Das Symbol Argument in der Kopie von Symbol Aufruf ist falsch. Es sollte so aussehen:

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