En CUDA, podemos lograr la transferencia de datos administrada por el kernel desde la memoria del host a la memoria compartida del dispositivo mediante el puntero del lado del dispositivo de la memoria del host. Me gusta esto:
int *a,*b,*c; // host pointers
int *dev_a, *dev_b, *dev_c; // device pointers to host memory
…
cudaHostGetDevicePointer(&dev_a, a, 0); // mem. copy to device not need now, but ptrs needed instead
cudaHostGetDevicePointer(&dev_b, b, 0);
cudaHostGetDevicePointer(&dev_c ,c, 0);
…
//kernel launch
add<<<B,T>>>(dev_a,dev_b,dev_c);
// dev_a, dev_b, dev_c are passed into kernel for kernel accessing host memory directly.
En el ejemplo anterior, el código del kernel puede acceder a la memoria del host a través de dev_a, dev_b y dev_c. Kernel puede utilizar estos punteros para mover los datos del host a la memoria compartida directamente sin transmitirlos por la memoria global.
¿Pero parece que es una misión imposible en OpenCL? (la memoria local en OpenCL es la contraparte de la memoria compartida en CUDA)
Respuestas
3 para la respuesta № 1Puedes encontrar exactamente la API idéntica en OpenCL.
Cómo funciona en CUDA:
De acuerdo a esta presentación y el documentación oficial.
La cotización del dinero sobre cudaHostGetDevicePointer :
Pasa el indicador de dispositivo de la memoria del host asignada asignada por cudaHostAlloc o registrado por cudaHostRegister.
CUDA cudaHostAlloc con cudaHostGetDevicePointer funciona exactamente como CL_MEM_ALLOC_HOST_PTR con MapBuffer Trabaja en OpenCL. Básicamente, si se trata de una GPU discreta, los resultados se almacenan en caché en el dispositivo y si se trata de una GPU discreta con memoria compartida con el host, utilizará la memoria directamente. Así que no hay una operación real de "copia cero" con GPU discreta en CUDA.
La función cudaHostGetDevicePointer no toma los punteros en bruto en mal estado, al igual que¿Cuál es la limitación en OpenCL. Desde el punto de vista de los usuarios de la API, esos dos son enfoques exactamente idénticos que permiten que la implementación realice optimizaciones prácticamente idénticas.
Con la GPU discreta, el puntero le da puntos aun área donde la GPU puede transferir material directamente a través de DMA. De lo contrario, el conductor tomaría su puntero, copiaría los datos al área DMA y luego iniciaría la transferencia.
Sin embargo en OpenCL2.0 que es explícitamente posible, dependiendo de las capacidades de sus dispositivos. Con el mejor intercambio de granularidad, puede usar punteros de host mal distribuidos al azar e incluso usar atomics con el host, por lo que incluso podría controlar dinámicamente el kernel desde el host mientras se está ejecutando.
http://www.khronos.org/registry/cl/specs/opencl-2.0.pdf
Consulte la página 162 para la especificación de memoria virtual compartida. Tenga en cuenta que cuando escribe kernels, incluso estos son solo unos __ punteros globales desde el punto de vista del kernel.