cuda - シンボルアドレスをシンボルにコピーします

Question

ホストからのシンボルアドレスへのシンボルポインタを設定したいのですが。何かのようなもの：

__device__ float * symbolPtr;   // address of symbol[3]
__device__ float symbol[5];

cudaGetSymbolAddress( &symbolPtr, symbol[3] );

これは、ホストへのデバイスアドレスを取得するためにのみ使用しました。ただし、シンボル変数を設定できないことは言及されていません。

別のアプローチは次のとおりです。

cudaMemcpytoSymbol(&symbolPtr, &symbol[3], sizeof(void*), 0, cudaMemcpyDeviceToDevice);

これらのどちらかがうまくいくとは思えません。これを行うためのより簡単な方法はありますか？

Answer 1

静的なアプローチは次のようになります。

__device__ float symbol[5];
__device__ float * symbolPtr= &symbol[3];

Answer 2

この方法は私にとってはうまくいきました。もっと良い方法があるかもしれません。

#include <stdio.h>
#define N 5
#define Q 3

#define cudaCheckErrors(msg) \
    do { \
        cudaError_t __err = cudaGetLastError(); \
        if (__err != cudaSuccess) { \
            fprintf(stderr, "Fatal error: %s (%s at %s:%d)\n", \
                msg, cudaGetErrorString(__err), \
                __FILE__, __LINE__); \
            fprintf(stderr, "*** FAILED - ABORTING\n"); \
            exit(1); \
        } \
    } while (0)

__device__ float *symbolPtr;
__device__ float symbol[N];

__global__ void mykernel(){

  for (int i=0; i<N; i++)
    symbol[i] = (float) i;
  printf("symbol[%d] = %f\n", Q, *symbolPtr);

}

int main(){

  float *temp;
  cudaGetSymbolAddress((void **)&temp, symbol);
  cudaCheckErrors("cudaGetSymbolAddress fail");
  temp += Q;
  cudaMemcpyToSymbol(symbolPtr, &temp, sizeof(float *));
  cudaCheckErrors("cudaMemcpyToSymbol fail");

  mykernel<<<1,1>>>();
  cudaDeviceSynchronize();
  cudaCheckErrors("kernel fail");


  return 0;
}