Trying to use global device memory

Hello:
I’m trying to use some __device__ memory that holds some useful values (read-only). So, starting from some code I already have, this is the overview

//BigInteger.h
#if CUDA_ENABLED == 1
 __device__ static BigInteger _cZERO;
 __device__ static BigInteger _cONE;
 __device__ static BigInteger _cTWO;
 __device__ static BigInteger _cTHREE;
 __device__ static BigInteger _cFOUR;
 __device__ static BigInteger _cFIVE;
 __device__ static BigInteger _cSIX;
 __device__ static BigInteger _cSEVEN;
 __device__ static BigInteger _cEIGHT;
 __device__ static BigInteger _cNINE;
 __device__ static BigInteger _cTEN;
 __device__ static BigInteger _cHUND;
 __device__ static BigInteger _cMIN;
#endif

static BigInteger _ZERO;
static BigInteger _ONE;
static BigInteger _TWO;
static BigInteger _THREE;
static BigInteger _FOUR;
static BigInteger _FIVE;
static BigInteger _SIX;
static BigInteger _SEVEN;
static BigInteger _EIGHT;
static BigInteger _NINE;
static BigInteger _TEN;
static BigInteger _HUND;
static BigInteger _MIN;

Then, on the code:

//BigInteger.c

void _BI_initialize() {
  int i = 0;

  //lo creamos manualmente ya que clean copia de _ZERO
  for (; i < MAX_LENGTH; i++)
    _ZERO.n[i] = 0;

  _ZERO.count = 0;
  _ZERO.k = 'i';

  newBI(&_ONE, "1", 0);
  newBI(&_TWO, "2", 0);
  newBI(&_THREE, "3", 0);
  newBI(&_FOUR, "4", 0);
  newBI(&_FIVE, "5", 0);
  newBI(&_SIX, "6", 0);
  newBI(&_SEVEN, "7", 0);
  newBI(&_EIGHT, "8", 0);
  newBI(&_NINE, "9", 0);
  newBI(&_TEN, "10", 0);
  newBI(&_HUND, "100", 0);
  newBI(&_MIN, "1", -1);

#if CUDA_ENABLED == 1
  //copy variables to device
  cudaMemcpyToSymbol(&_cZERO, &_ZERO, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cONE, &_ONE, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cTWO, &_TWO, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cTHREE, &_THREE, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cFOUR, &_FOUR, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cFIVE, &_FIVE, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cSIX, &_SIX, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cSEVEN, &_SEVEN, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cEIGHT, &_EIGHT, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cNINE, &_NINE, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cTEN, &_TEN, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cHUND, &_HUND, sizeof(BigInteger), cudaMemcpyHostToDevice);
  cudaMemcpyToSymbol(&_cMIN, &_MIN, sizeof(BigInteger), cudaMemcpyHostToDevice);
#endif
}

It compiles without problem, then when I run the code I checked

  • sizeof(BigInteger) = sizeof(_cZERO) = sizeof(_ZERO) = 4104
  • After cudaMemcpyToSymbol() any _device_ variable has been changed (seen on the VisualStudio debug panel).

What am I missing?

Thanks.

What problem do you have with your code? What is not working correctly? Can you show a minimal complete code example?

Hi:
Sorry for late response… I was checking on the Internet and I was facing two issues

  1. Invalid PTX (outdated driver)
  2. Wrong codification
cudaMemcpyToSymbol(&_cMIN, &_MIN, sizeof(BigInteger), cudaMemcpyHostToDevice);

needs to be

cudaMemcpyToSymbol(_cMIN, &_MIN, sizeof(BigInteger), 0, cudaMemcpyHostToDevice);

So now problem is solved :)