Fallo de segmentación de cudaMemcpy

Estuve atormentado por este error durante bastante tiempo, así que decidí publicarlo aquí.

Esta falla de segmentación ocurrió cuando se llama a cudaMemcpy:

CurrentGrid->cdata[i] = new float[size]; cudaMemcpy(CurrentGrid->cdata[i], Grid_dev->cdata[i], size*sizeof(float),\ cudaMemcpyDeviceToHost); 

CurrentGrid y Grid_dev son un puntero a un objeto de clase de grid en el host y el dispositivo respectivamente y i = 0 en este contexto. El cdata miembro de cdata es una matriz de puntero de tipo flotante. Para la depuración, justo antes de esta llamada cudaMemcpy imprimí el valor de cada elemento de Grid_Dev->cdata[i] , la dirección de CurrentGrid->cdata[i] y Grid_dev->cdata[i] y el valor de size , que todo se ve bien Pero todavía termina con “Error de segmentación (núcleo volcado)”, que es el único mensaje de error. cuda-memcheck solo dio “el proceso no finalizó exitosamente”. No puedo usar cuda-gdb en este momento. ¿Alguna sugerencia sobre dónde ir?

ACTUALIZACIÓN : Parece que ahora he resuelto este problema mediante cudaMalloc otro puntero flotante A en el dispositivo y cudaMemcpy el valor de Grid_dev-> cdata [i] a A, y luego cudaMemcpy A al host. Entonces, el segmento de código escrito arriba se convierte en:

 float * A; cudaMalloc((void**)&A, sizeof(float)); ... ... cudaMemcpy(&A, &(Grid_dev->cdata[i]), sizeof(float *), cudaMemcpyDeviceToHost); CurrentGrid->cdata[i] = new float[size]; cudaMemcpy(CurrentGrid->cdata[i], A, size*sizeof(float), cudaMemcpyDeviceToHost); 

Hice esto porque valgrind apareció “lectura inválida de tamaño 8”, que pensé que se refería a Grid_dev->cdata[i] . Lo comprobé nuevamente con gdb, imprimiendo el valor de Grid_dev->cdata[i] como NULL. Así que supongo que no puedo desreferenciar directamente el puntero del dispositivo incluso en esta llamada cudaMemcpy. Pero por qué ? De acuerdo con el comentario al final de este hilo , deberíamos poder desreferenciar el puntero del dispositivo en la función cudaMemcpy.

Además, no conozco el mecanismo subyacente de cómo funcionan cudaMalloc y cudaMemcpy, pero creo que con un puntero cudaMalloc, digamos A aquí, asignamos este puntero para señalar una determinada dirección en el dispositivo. Y por cudaMemcpy el Grid_dev->cdata[i] a A como en el código modificado anterior, reasignamos el puntero A para apuntar a la matriz. Entonces, ¿no perdemos la pista de la dirección anterior a la que A señaló cuando está cudaMalloced? ¿Podría esto causar pérdida de memoria o algo así? En caso afirmativo, ¿cómo debo solucionar esta situación de forma adecuada? ¡Gracias!

Como referencia, coloque el código de la función completa en la que ocurrió este error a continuación.

¡Muchas gracias!

 __global__ void Print(grid *, int); __global__ void Printcell(grid *, int); void CopyDataToHost(param_t p, grid * CurrentGrid, grid * Grid_dev){ cudaMemcpy(CurrentGrid, Grid_dev, sizeof(grid), cudaMemcpyDeviceToHost); #if DEBUG_DEV cudaCheckErrors("cudaMemcpy1 error"); #endif printf("\nBefore copy cell data\n"); Print<<>>(Grid_dev, 0); //Print out some Grid_dev information for cudaDeviceSynchronize(); //debug int NumberOfBaryonFields = CurrentGrid->ReturnNumberOfBaryonFields(); int size = CurrentGrid->ReturnSize(); int vsize = CurrentGrid->ReturnVSize(); CurrentGrid->FieldType = NULL; CurrentGrid->FieldType = new int[NumberOfBaryonFields]; printf("CurrentGrid size is %d\n", size); for( int i = 0; i 

cdata[i] = NULL; CurrentGrid->vdata[i] = NULL; CurrentGrid->cdata[i] = new float[size]; CurrentGrid->vdata[i] = new float[vsize]; Printcell<<>>(Grid_dev, i);//Print out element value of Grid_dev->cdata[i] cudaDeviceSynchronize(); cudaMemcpy(CurrentGrid->cdata[i], Grid_dev->cdata[i], size*sizeof(float),\ cudaMemcpyDeviceToHost); //where error occurs #if DEBUG_DEV cudaCheckErrors("cudaMemcpy2 error"); #endif printf("\nAfter copy cell data\n"); Print<<>>(Grid_dev, i); cudaDeviceSynchronize(); cudaMemcpy(CurrentGrid->vdata[i], Grid_dev->vdata[i], vsize*sizeof(float),\ cudaMemcpyDeviceToHost); #if DEBUG_DEV cudaCheckErrors("cudaMemcpy3 error"); #endif } cudaMemcpy(CurrentGrid->FieldType, Grid_dev->FieldType,\ NumberOfBaryonFields*sizeof(int), cudaMemcpyDeviceToHost); #if DEBUG_DEV cudaCheckErrors("cudaMemcpy4 error"); #endif }

EDITAR: aquí está la información de valgrind, desde la cual bash rastrear dónde ocurrió la pérdida de memoria.

 ==19340== Warning: set address range perms: large range [0x800000000, 0xd00000000) (noaccess) ==19340== Warning: set address range perms: large range [0x200000000, 0x400000000) (noaccess) ==19340== Invalid read of size 8 ==19340== at 0x402C79: CopyDataToHost(param_t, grid*, grid*) (CheckDevice.cu:48) ==19340== by 0x403646: CheckDevice(param_t, grid*, grid*) (CheckDevice.cu:186) ==19340== by 0x40A6CD: main (Transport.cu:81) ==19340== Address 0x2003000c0 is not stack'd, malloc'd or (recently) free'd ==19340== ==19340== ==19340== Process terminating with default action of signal 11 (SIGSEGV) ==19340== Bad permissions for mapped region at address 0x2003000C0 ==19340== at 0x402C79: CopyDataToHost(param_t, grid*, grid*) (CheckDevice.cu:48) ==19340== by 0x403646: CheckDevice(param_t, grid*, grid*) (CheckDevice.cu:186) ==19340== by 0x40A6CD: main (Transport.cu:81) ==19340== ==19340== HEAP SUMMARY: ==19340== in use at exit: 2,611,365 bytes in 5,017 blocks ==19340== total heap usage: 5,879 allocs, 862 frees, 4,332,278 bytes allocated ==19340== ==19340== LEAK SUMMARY: ==19340== definitely lost: 0 bytes in 0 blocks ==19340== indirectly lost: 0 bytes in 0 blocks ==19340== possibly lost: 37,416 bytes in 274 blocks ==19340== still reachable: 2,573,949 bytes in 4,743 blocks ==19340== suppressed: 0 bytes in 0 blocks ==19340== Rerun with --leak-check=full to see details of leaked memory ==19340== ==19340== For counts of detected and suppressed errors, rerun with: -v ==19340== ERROR SUMMARY: 1 errors from 1 contexts (suppressed: 2 from 2) 

Creo que sé cuál es el problema, pero para confirmarlo, sería útil ver el código que está utilizando para configurar las clases Grid_dev en el dispositivo.

Cuando una clase u otra estructura de datos se va a usar en el dispositivo, y esa clase tiene punteros que hacen referencia a otros objetos o almacenamientos intermedios en la memoria (presumiblemente en la memoria del dispositivo, para una clase que se usará en el dispositivo), el proceso de hacer que esta clase de nivel superior sea utilizable en el dispositivo se vuelve más complicado.

Supongamos que tengo una clase como esta:

 class myclass{ int myval; int *myptr; } 

Podría crear una instancia de la clase anterior en el host, y luego malloc una matriz de int y asignar ese puntero a myptr , y todo estaría bien. Para que esta clase se pueda utilizar en el dispositivo y solo en el dispositivo, el proceso podría ser similar. Yo podría:

  1. cudaMalloc un puntero a la memoria del dispositivo que contendrá myclass
  2. (opcionalmente) copie un objeto instanciado de myclass en el host al puntero del dispositivo desde el paso 1 usando cudaMemcpy
  3. en el dispositivo, use malloc o new para asignar el almacenamiento del dispositivo para myptr

La secuencia anterior está bien si nunca quiero acceder al almacenamiento asignado para myptr en el host. Pero si quiero que ese almacenamiento sea visible desde el host, necesito una secuencia diferente:

  1. cudaMalloc un puntero a la memoria del dispositivo que contendrá myclass , llamémoslo mydevobj
  2. (opcionalmente) copie un objeto instanciado de myclass en el host al puntero del dispositivo mydevobj del paso 1 usando cudaMemcpy
  3. Cree un puntero int por separado en el host, llamémoslo myhostptr
  4. cudaMalloc int storage en el dispositivo para myhostptr
  5. cudaMemcpy el valor del puntero de myhostptr desde el host al puntero del dispositivo &(mydevobj->myptr)

Después de eso, puede cudaMemcpy los datos apuntados por el puntero incrustado myptr a la región asignada (a través de cudaMalloc ) en myhostptr

Tenga en cuenta que en el paso 5, porque estoy tomando la dirección de esta ubicación del puntero, esta operación cudaMemcpy solo requiere el puntero mydevobj en el host, que es válido en una operación cudaMemcpy (solamente).

El valor del puntero del dispositivo myint se configurará correctamente para realizar las operaciones que está intentando hacer. Si luego desea utilizar los datos de myint desde y hacia myint para el host, use el puntero myhostptr en cualquier llamada a cudaMemcpy, no en mydevobj->myptr . Si tratamos de usar mydevobj->myptr , sería necesario desreferenciar mydevobj y luego usarlo para recuperar el puntero que está almacenado en myptr , y luego usar ese puntero como la ubicación de copia hacia / desde. Esto no es aceptable en el código de host. Si intenta hacerlo, obtendrá un error seg. (Tenga en cuenta que a modo de analogía, mi mydevobj es como su Grid_dev y mi myptr es como su cdata )

En general, es un concepto que requiere una reflexión cuidadosa la primera vez que se encuentra con él, por lo que preguntas como esta aparecen con cierta frecuencia en SO. Es posible que desee estudiar algunas de estas preguntas para ver ejemplos de código (ya que no ha proporcionado su código que configura Grid_dev ):

  1. Ejemplo 1
  2. ejemplo 2
  3. ejemplo 3