统一的虚拟地址空间:
当应用程序以64位进程运行时,单个地址空间用于主机以及计算能力2.0及更高版本的所有设备。 所有通过CUDA API调用进行的主机内存分配以及受支持设备上的所有设备内存分配均在此虚拟地址范围内。 作为结果:
- 可以使用cudaPointerGetAttributes()从指针的值中确定通过CUDA分配的主机上的任何内存或使用统一地址空间的任何设备上的内存位置。
- 在使用统一地址空间的任何设备的内存中进行复制或复制时,可以将cudaMemcpy *()的cudaMemcpyKind参数设置为cudaMemcpyDefault,以确定指针的位置。 这也适用于未通过CUDA分配的主机指针,只要当前设备使用统一寻址。
- 通过cudaHostAlloc()的分配在所有使用统一地址空间的设备上是自动可移植的,并且由cudaHostAlloc()返回的指针可以直接从在这些设备上运行的内核中使用(即,不存在 需要通过cudaHostGetDevicePointer()获取设备指针,如Mapped Memory中所述
应用程序可以通过检查unifiedAddressing设备属性(请参阅设备枚举)是否等于1来查询统一地址空间是否用于特定设备。
进程间通信:
由主机线程创建的任何设备内存指针或事件句柄都可以被同一进程内的任何其他线程直接引用。 然而,在这个过程之外它是无效的,因此不能被属于不同进程的线程直接引用。
要跨进程共享设备内存指针和事件,应用程序必须使用参考手册中详细描述的进程间通信API。 IPC API仅支持Linux上的64位进程以及计算能力为2.0或更高的设备。
使用这个API,应用程序可以使用cudaIpcGetMemHandle()来获得给定设备内存指针的IPC句柄,并使用标准IPC机制(例如进程间共享内存或文件)将其传递给另一个进程,并使用cudaIpcOpenMemHandle()来检索设备 来自IPC句柄的指针,这是另一个进程中有效的指针。 事件句柄可以使用类似的入口点共享。
使用IPC API的一个例子是单个主进程生成一批输入数据,使数据可用于多个从进程而无需再生或复制。
错误检查:
所有运行时函数都会返回一个错误代码,但对于异步函数(请参见异步并发执行),此错误代码不可能报告设备上可能发生的任何异步错误,因为该函数在设备完成任务之前返回; 错误代码仅报告执行任务之前在主机上发生的错误,通常与参数验证有关; 如果发生异步错误,则会由一些后续不相关的运行时函数调用报告。
因此,在调用一个异步函数之后检查异步错误的唯一方法就是在调用之后通过调用cudaDeviceSynchronize()(或者使用异步并发执行中描述的任何其他同步机制)同步并检查由cudaDeviceSynchronize返回的错误代码()
运行时为每个主机线程维护一个错误变量,该主机线程被初始化为cudaSuccess,并且每次发生错误时都会被错误代码覆盖(不管是参数验证错误还是异步错误)。 cudaPeekAtLastError()返回这个变量。 cudaGetLastError()返回这个变量并将其重置为cudaSuccess。
内核启动不返回任何错误代码,所以cudaPeekAtLastError()或
必须在内核启动后调用cudaGetLastError()才能检索任何启动前错误。 为了确保由cudaPeekAtLastError()或cudaGetLastError()返回的任何错误不是来自内核启动之前的调用,必须确保运行时错误变量在内核启动之前设置为cudaSuccess,例如通过调用 内核启动之前的cudaGetLastError()。 内核启动是异步的,因此为了检查异步错误,应用程序必须在内核启动和对cudaPeekAtLastError()或cudaGetLastError()的调用之间进行同步。
请注意,可能由cudaStreamQuery()和cudaEventQuery()返回的cudaErrorNotReady不被视为错误,因此不会被cudaPeekAtLastError()或cudaGetLastError()报告。
调用堆栈:
在计算能力为2.x或更高的设备上,可以使用cudaDeviceGetLimit()查询调用堆栈的大小,并使用cudaDeviceSetLimit()
当调用堆栈溢出时,如果应用程序通过CUDA调试器(cuda-gdb,Nsight)或未指定的启动错误运行,那么内核调用会失败并出现堆栈溢出错误。