pointers - sirve - puntero malloc

Copiar a la memoria global asignada por malloc()?

Question

La guía de programación de CUDA indica que "La memoria asignada mediante malloc() se puede copiar utilizando el tiempo de ejecución (es decir, llamando a cualquiera de las funciones de la memoria de copia desde la memoria del dispositivo)", pero de alguna manera tengo problemas para reproducir esta funcionalidad. Código:

#include <cstdio>
__device__ int* p;
__global__ void allocate_p() {
p = (int*) malloc(10);
printf("p = %p (seen by GPU)/n", p);
}
int main() {
cudaError_t err;
int* localp = (int*) malloc(10);
allocate_p<<<1,1>>>();
cudaDeviceSynchronize();
//Getting pointer to device-allocated memory
int* tmpp = NULL;
cudaMemcpyFromSymbol(&tmpp, p, 4);
printf("p = %p (seen by CPU)/n", tmpp);
//cudaMalloc((void**)&tmpp, 40);
err = cudaMemcpy(tmpp, localp, 40, cudaMemcpyHostToDevice);
cudaDeviceSynchronize();
printf(" err:%i %s", (int)err, cudaGetErrorString(err));
delete localp;
return 0;
}

cuelga con salida:

p = 0x601f920 (seen by GPU)
p = 0x601f920 (seen by CPU)
err:11 invalid argument

Supongo que el anfitrión ve la dirección apropiada en el dispositivo, pero de alguna manera no le gusta que provenga de malloc() .

Si cudaMalloc((void**)&np, 40); antes por cudaMalloc((void**)&np, 40); y luego pase el puntero np como argumento a kernel allocate_p , donde se asignará a p (en lugar de malloc() ), y luego el código se ejecutará correctamente.

¿Qué estoy haciendo mal / cómo usamos la memoria de dispositivo asignada malloc() en las funciones del lado del servidor?

Answer 1

Por lo que sé, no es posible copiar la memoria del montón en tiempo de ejecución utilizando las funciones de la API de host. Ciertamente no fue posible en CUDA 4.x y el candidato de lanzamiento de CUDA 5.0 no ha cambiado esto. La única solución que puedo ofrecer es usar un kernel para "recopilar" los resultados finales y rellenarlos en un buffer de transferencia de dispositivo o en una memoria de cero copia a la que se puede acceder a través de la API o directamente desde el host. Puede ver un ejemplo de este enfoque en esta respuesta y otra pregunta donde Mark Harris de NVIDIA confirmó que esto es una limitación de la (actual) implementación actual en el tiempo de ejecución de CUDA.