title: How do I convert a std::vector<thrust::device_vector<int>> to int**?
body: <p>I am working on an application in which previous processing has produced a (shortish but variable-length) <code>std::vector</code> of (big) <code>thrust::device_vector</code>s, each with the same length (but that length is also variable). I need to convert it to a raw pointer on the device to pass it to a cuda kernel.</p>
  <p>I did the process below, which as far as I can see should leave <code>rawNumberSquare</code> as a pointer on the device, with <code>rawNumberSquare[0]</code> and <code>rawNumberSquare[1]</code> each containing a pointer to <code>numberSquareOnDevice[0][0]</code> and <code>numberSquareOnDevice[1][0]</code> respectively. So, it seems to me that <code>rawNumberSquare[i][j]</code> (i,j = 0,1) are all locations assigned by this program and legal to access.</p>
  <p>However when a kernel tries to access these location the values are wrong and the program crashes with an illegal memory access.</p>
  <pre><code>#include &quot;cuda_runtime.h&quot;
  #include &quot;device_launch_parameters.h&quot;
  #include &lt;stdio.h&gt;
  #include&lt;vector&gt;
  #include&lt;thrust/device_vector.h&gt;
  
  __global__ void talkKernel(  int ** in,  int dimension)
  {
      int index = threadIdx.x;
      for (int coord = 0; coord &lt; dimension; ++coord)
          printf(&quot;in[%d][%d] = %d\n&quot;, coord, index, in[coord][index]);       
  }
  
  int main()
  {
      //print out name of GPU in case it is helpful
      int deviceNumber;
      cudaGetDevice(&amp;deviceNumber);
      cudaDeviceProp prop;
      cudaGetDeviceProperties(&amp;prop, deviceNumber);
      std::cout &lt;&lt; prop.name &lt;&lt; &quot;\n&quot;;
      //make a std::vector of std::vectors of ints
      std::vector&lt;std::vector&lt;int&gt;&gt; numberSquareOnHost{ {1,2}, {3,4} };
      //copy the values of each vector to the device
      std::vector&lt;thrust::device_vector&lt;int&gt;&gt; numberSquareDevice;
      for (auto&amp; vector : numberSquareOnHost)
          numberSquareDevice.push_back(thrust::device_vector&lt;int&gt;(vector));
      //copy the raw pointers to start of the device vectors to a std::vector
      std::vector&lt;int*&gt; halfRawNumberSquareOnHost(2);
      for ( int i = 0; i &lt; 2 ; ++i)
          halfRawNumberSquareOnHost[i] = (thrust::raw_pointer_cast(numberSquareOnHost[i].data()));
      //copy the raw pointers ot the device
      thrust::device_vector&lt;int*&gt; halfRawNumberSquareOnDevice(halfRawNumberSquareOnHost);
      //get raw pointer (on the device) to the raw pointers (on the device)
      int** rawNumberSquare = thrust::raw_pointer_cast(halfRawNumberSquareOnDevice.data());
      //call the kernel
      talkKernel &lt;&lt;&lt;1,2 &gt;&gt;&gt; ( rawNumberSquare, 2);
      cudaDeviceSynchronize();
      //ask what's up'
      std::cout &lt;&lt; cudaGetErrorString(cudaGetLastError()) &lt;&lt; &quot;\n&quot;;
      return 0;
  
     /*output:
     * Quadro M2200
      in[0][0] = 0
      in[0][1] = 0
      in[1][0] = 0
      in[1][1] = 0
      an illegal memory access was encountered
  
      ...\vectorOfVectors.exe (process 6428) exited with code -1073740791.
          */
  }
  </code></pre>
  <p>I also tried all such as allocating a host pointer to (raw device) <code>int*</code> with <code>new</code> rather than using the <code>std::vector&lt;int*&gt; halfRawNumberSquareOnHost</code> and allocating the device <code>int** rawSquareOnDevice</code> with <code>cudaMalloc</code> (and filling it with <code>cudaMemcpy</code>). This didn't make a difference.</p>