1

I have made some helper functions for doing operations with CUDA __constant__ pointers (allocation, copyToSymbol, copyFromSymbol, etc). I also have error checking in place as suggested by talonmies here. Here is a basic working example:

#include <cstdio>
#include <cuda_runtime.h>

__constant__ float* d_A;

__host__ void cudaAssert(cudaError_t code,
                         char* file,
                         int line,
                         bool abort=true) {
  if (code != cudaSuccess) {
    fprintf(stderr, "CUDA Error: %s in %s at line %d\n",
           cudaGetErrorString(code), file, line);
    if (abort) {
      exit(code);
    }   
  }
}

#define cudaTry(ans) { cudaAssert((ans), __FILE__, __LINE__); }

template<typename T>
void allocateCudaConstant(T* &d_ptr,
                          size_t size) {
  size_t memsize = size * sizeof(T);
  void* ptr;
  cudaTry(cudaMalloc((void**) &ptr, memsize));
  cudaTry(cudaMemset(ptr, 0, memsize));
  cudaTry(cudaMemcpyToSymbol(d_ptr, &ptr, sizeof(ptr),
                             0, cudaMemcpyHostToDevice));
}

int main() {
  size_t size = 16; 
  allocateCudaConstant<float>(d_A, size);
  return 0;
}

When I compile this with nvcc, I get the following warning:

In file included from tmpxft_0000a3e8_00000000-3_example.cudafe1.stub.c:2:
example.cu: In function ‘void allocateCudaConstant(T*&, size_t) [with T = float]’:
example.cu:35:   instantiated from here
example.cu:29: warning: deprecated conversion from string constant to ‘char*’

I understand what the warning means, but I can't for the life of me figure out where it is coming from. If I don't make allocateCudaConstant a template function, I don't get the warning. If I don't wrap cudaMemcpyToSymbol in cudaTry, I also don't get the warning. I know it is just a warning, and if I compile with -Wno-write-strings I can suppress the warning. The code runs fine but I don't want to get in the habit of ignoring warnings and if I suppress the warnings, I may hide other issues that need to be addressed.

So, can anyone help me figure out where the warning is coming from and how I can suppress it?

Community
  • 1
  • 1
joelmeans
  • 122
  • 1
  • 1
  • 8
  • "I know it is just a warning" - in C++11, it should be an error. The dodgy conversion is no longer deprecated, it's not allowed at all. – Mike Seymour Nov 12 '13 at 15:40
  • @MikeSeymour Another reason to make sure I take care of it now… Assuming `nvcc` adds C++11 compatibility sometime soon. – joelmeans Nov 12 '13 at 15:47

1 Answers1

5

Change char* file to const char* file in the declaration of cudaAssert. You don't need to modify the string, so you should not ask for a modifyable string.

Kerrek SB
  • 464,522
  • 92
  • 875
  • 1,084
  • Thanks. That fixed the issue. Care to give an explanation of why this error pops up for the template function but not in the case where I explicitly make it a `float` or `double` function? – joelmeans Nov 12 '13 at 15:36
  • @joelmeans: well, you were passing an unmodifiable string to a function that expects a modifiable one. – Kerrek SB Nov 12 '13 at 15:38
  • Perhaps adding the template causes nvcc to use g++ while without it, gcc will suffice and the conversion is not deprecated in C? Just spitballing here. – joelmeans Nov 12 '13 at 15:43
  • @joelmeans: That's right, C and C++ are different languages with different rules, especially pertaining to string literals. – Kerrek SB Nov 12 '13 at 16:02