Uso de memoria constante CUDA en múltiples archivos fuente mostrando diferentes comportamientos en cuda-11.2 y cuda-11.4.
Mínimo reproducible:
kernel.cu:
<h1>#include <stdio.h></h1>
<p><strong>constant</strong> int N_GPU;</p>
<p>void wrapper_fn(int *ptr)
{
cudaMemcpyToSymbol(N_GPU, ptr, sizeof(int), cudaMemcpyDeviceToDevice);
}</p>
<p><strong>global</strong> void printKernel() {</p>
<pre><code>printf("N = %d; \n", N_GPU);
</code></pre>
<p>}
<code>
driver.cu:
</code>c</p>
<h1>#include "cuda_runtime.h"</h1>
<h1>#include <stdio.h></h1>
<p>void wrapper_fn(int*);
<strong>global</strong> void printKernel();</p>
<p>int main()
{
int N = 10;
int* d_N_ptr;
cudaMalloc(&d_N_ptr, sizeof(int));
cudaMemcpy(d_N_ptr, &N, sizeof(int), cudaMemcpyDefault);</p>
<pre><code>wrapper_fn(d_N_ptr);
printKernel <<<1, 1 >>>();
cudaPeekAtLastError();
cudaDeviceSynchronize();
return 0;
</code></pre>
<p>}
```</p>
Tanto en cuda-11.4 como en cuda-11.2, ejecutando <code>nvcc kernel.cu driver.cu</code> se compila. La salida esperada (es decir, <code>N = 10;</code>) solo se ve en el 11.2 y no en el 11.4.
Al ejecutar <code>cuda-gdb</code> en el 11.4, obtengo lo siguiente:
</code></pre>
...
[New Thread 0x7fffee240700 (LWP 54339)]
warning: Error detectado en la API de Cuda: cudaMalloc devolvió (0xde)</p>
warning: Error detectado en la API de Cuda: cudaMemcpy devolvió (0xde)
warning: Error detectado en la API de Cuda: cudaMemcpyToSymbol devolvió (0xde)
warning: Error detectado en la API de Cuda: cudaLaunchKernel devolvió (0xde)
warning: Error detectado en la API de Cuda: cudaPeekAtLastError devolvió (0xde)
warning: Error detectado en la API de Cuda: cudaDeviceSynchronize devolvió (0xde)
[Thread 0x7fffee240700 (LWP 54339) salió]
...
```
¿Hay alguna bandera de nvcc particular que me falte y sea importante en el 11.4? ¿O cambios particulares en la API que me esté perdiendo? ¡Gracias de antemano!
davy.ai
Parece que el problema es causado por el uso de
cudaMemcpyDefaulten la llamadacudaMemcpyen el archivodriver.cu.A partir de CUDA 11.3, el uso de
cudaMemcpyDefaultya no es compatible y debe reemplazarse con uno de los tipos específicos de cudaMemcpy (es decir,cudaMemcpyHostToHost,cudaMemcpyHostToDevice,cudaMemcpyDeviceToHostocudaMemcpyDeviceToDevice).Para solucionar el problema, reemplace
cudaMemcpyDefaultconcudaMemcpyHostToDeviceen el archivodriver.cu, de la siguiente manera:Después de realizar este cambio, el código debería funcionar como se esperaba tanto en CUDA 11.4 como en CUDA 11.2.