cudaInvalidSymbol из-за cudaMemcpyToSymbol на структуру

Question

Вот минимальный пример, который бросает cudaInvalidSymbol на sync_and_check("cudaMemcpyToSymbol"), составленное с использованием Cuda инструменты компиляции, выпуск 6.5, V6.5.12:

#include<stdio.h> 

__inline __host__ void gpuAssert(cudaError_t code, const char* command, const char *file, int line, 
       bool abort=true) 
{ 
    if (code != cudaSuccess) 
    { 
     fprintf(stderr,"Error while calling %s in %s line %d: %s\n", command, 
      file, line, cudaGetErrorString(code)); 
     if (abort) exit(code); 
    } 
} 

#define sync_and_check(command) { gpuAssert(cudaGetLastError(), command, __FILE__, __LINE__); } 

struct S 
{ 
    float one; 
}; 

__device__ __constant__ S d_s; 

__global__ void kernel(float* f) 
{ 
    (*f) = d_s.one; 
} 

int main() 
{ 
    cudaGetLastError(); 

    S s; 
    s.one = 1.f; 
    cudaMemcpyToSymbol(&d_s, &s, sizeof(S)); 
    sync_and_check("cudaMemcpyToSymbol"); 

    float* d_f; 
    cudaMalloc(&d_f, sizeof(float)); 
    sync_and_check("cudaMalloc"); 

    dim3 dimGrid(1), dimBlock(32); 
    kernel<<<dimGrid, dimBlock>>>(d_f); 
    sync_and_check("kernel"); 

    cudaFree(d_f); 
    sync_and_check("cudaFree"); 
} 

Answer 1

Проблема заключается в ссылки. Правильный вызов является

cudaMemcpyToSymbol(d_s, &s, sizeof(S)); 

я был смущен, потому что страница человек говорит

cudaError_t cudaMemcpyToSymbol (const void * symbol, const void * src, 
    size_t count, size_t offset = 0, enum cudaMemcpyKind kind = 
    cudaMemcpyHostToDevice) 
    Copies count bytes from the memory area pointed to by src to the memory 
    area pointed to by offset bytes from the start of symbol symbol. The 
    memory areas may not overlap. symbol is a variable that resides in 
    global or constant memory space. kind can be either 
    cudaMemcpyHostToDevice or cudaMemcpyDeviceToDevice. 

    Parameters: 
     symbol - Device symbol address 

Все, что они подразумевают под «символом» :) Копирование на структуру элементов

cudaMemcpyToSymbol(d_s.one, &(s.one), sizeof(S)); 

работает так же хорошо ,