title: Cublas not working within kernel once compiled to cubin using -G flag with nvcc
body: <p>I have a CUDA kernel that looks like the following:</p>
  
  <pre><code>#include &lt;cublas_v2.h&gt;
  #include &lt;math_constants.h&gt;
  #include &lt;stdio.h&gt;
  
  
  extern "C" {
  
      __device__ float ONE = 1.0f;
      __device__ float M_ONE = -1.0f;
      __device__ float ZERO = 0.0f;
  
      __global__ void kernel(float *W, float *input, int i, float *output, int o) {
          int idx = blockIdx.x*blockDim.x+threadIdx.x;
          cublasHandle_t cnpHandle;
  
          if(idx == 0) {
              cublasCreate(&amp;cnpHandle);
              cublasStatus_t s = cublasSgemv(cnpHandle, CUBLAS_OP_N, o, i, &amp;ONE, W, o, input, 1, &amp;ZERO, output, 1);
              printf("status %d\n", s);
              cudaError_t e = cudaDeviceSynchronize();
              printf("sync %d\n", e);
          }
  
      }
  
  }
  </code></pre>
  
  <p>The host code:</p>
  
  <pre><code>#include &lt;iostream&gt;
  #include &lt;numeric&gt;
  #include &lt;stdlib.h&gt;
  #include &lt;cstring&gt;
  #include &lt;cuda_runtime.h&gt;
  #include &lt;cuda.h&gt;
  #include &lt;cublas_v2.h&gt;
  
  extern "C" {
      __global__ void kernel(float *W, float *input, int i, float *output, int o);
  }
  
  #define gpuErrchk(ans) { gpuAssert((ans), __FILE__, __LINE__); }
  inline void gpuAssert(cudaError_t code, const char *file, int line, bool abort=true)
  {
     if (code != cudaSuccess)
     {
        fprintf(stderr,"GPUassert: %s %s %d\n", cudaGetErrorString(code), file, line);
        if (abort) exit(code);
     }
  }
  
  int main(int argc, char* argv[])
  {
  
      cuInit(0);
      CUcontext pctx;
      CUdevice dev;
      cuDeviceGet(&amp;dev, 0);
      cuCtxCreate(&amp;pctx, 0, dev);
  
      CUmodule module;
      CUresult t = cuModuleLoad(&amp;module, "pathto/src/minimalKernel.cubin");
  
      CUfunction function;
      CUresult r = cuModuleGetFunction(&amp;function, module, "kernel");
  
      float *W = new float[2];
      W[0] = 0.1f;
      W[1] = 0.1f;
      float *input = new float[2];
      input[0] = 0.1f;
      input[1] = 0.1f;
      float *out = new float[1];
      out[0] = 0.0f;
  
      int i = 2;
      int o = 1;
  
      float *d_W;
      float *d_input;
      float *d_out;
      cudaMalloc((void**)&amp;d_W, 2*sizeof(float));
      cudaMalloc((void**)&amp;d_input, 2*sizeof(float));
      cudaMalloc((void**)&amp;d_out, sizeof(float));
      cudaMemcpy(d_W, W, 2*sizeof(float), cudaMemcpyHostToDevice);
      cudaMemcpy(d_input, input, 2*sizeof(float), cudaMemcpyHostToDevice);
      cudaMemcpy(d_out, out, sizeof(float), cudaMemcpyHostToDevice);
      //kernel&lt;&lt;&lt;1, 2&gt;&gt;&gt;(d_W, d_input, i, d_out, o);
  
      //cudaMemcpy(out, d_out, sizeof(float), cudaMemcpyDeviceToHost);
  
      //std::cout&lt;&lt;"out:"&lt;&lt;out[0]&lt;&lt;std::endl;
  
      void * kernelParams[] { &amp;d_W, &amp;d_input, &amp;i, &amp;d_out, &amp;o };
  
      CUresult k = cuLaunchKernel(function, 1, 1, 1, 2, 1, 1, 0, 0, (void**)kernelParams, 0);
      gpuErrchk( cudaPeekAtLastError() );
      gpuErrchk( cudaDeviceSynchronize() );
      cudaMemcpy(out, d_out, sizeof(float), cudaMemcpyDeviceToHost);
  
      std::cout&lt;&lt;"out:"&lt;&lt;out[0]&lt;&lt;std::endl;
  
  
  }
  </code></pre>
  
  <p>When this kernel runs inline <code>kernel&lt;&lt;&lt;1,2&gt;&gt;&gt;()</code>, built and linked (within eclipse Nsight), the kernel runs completely fine and <code>out</code> returns <code>0.02</code>
  as expected.</p>
  
  <p>If I compile the kernel into a .cubin using <code>-G</code> (generate device debugging symbols), the cublas function never runs, and the <code>out</code> is always <code>0.0</code></p>
  
  <p>I can put breakpoints in when the .cubin is running and I can see the data is correct going into the cublas function, but it looks like the cublas function never runs at all. The cublas function also always is returning 0 <code>CUDA_SUCCESS</code>. Importantly this ONLY happens when running this from a .cubin</p>
  
  <p>To compile to a cubin I am using with the <code>-G</code>:</p>
  
  <pre><code>nvcc -G -cubin -arch=sm_52 --device-c kernel.cu -o kernel.cubin -dlink -lcudadevrt -lcublas_device
  </code></pre>
  
  <p>which returns no errors. </p>
  
  <p>Why would the cublas functions within the .cubin stop working if the <code>-G</code> option is added?</p>
  
  <p>CUDA 7.0
  linux 14.04 x64
  980GTX </p>