深度学习训练中GPU内存的专用与共享机制
深度学习训练中GPU内存的专用与共享机制
在深度学习和GPU编程中，理解GPU内存的访问机制对于优化程序性能至关重要。本文将深入探讨在不同计算能力的GPU设备上，CPU和GPU对统一内存的访问机制，以及如何确保数据的一致性和并发性。
一致性和并发性
在计算能力低于6.x的设备上同时访问托管内存是不可能的，因为如果CPU在GPU内核处于活动状态时访问统一内存分配，则无法保证一致性。 但是，支持操作系统的计算能力6.x设备允许CPU和GPU通过新的页面错误机制同时访问统一内存分配。 程序可以通过检查新的concurrentManagedAccess属性来查询设备是否支持对托管内存的并发访问。 请注意，与任何并行应用程序一样，开发人员需要确保正确的同步以避免处理器之间的数据危害。
GPU独占访问管理内存
为了确保6.x之前的GPU架构的一致性，Unified Memory编程模型在数据访问方面存在约束条件，而CPU和GPU同时执行。 实际上，在任何内核操作正在执行时，GPU都可独占访问所有受管数据，而不管特定内核是否正在使用该数据。 当托管数据与cudaMemcpy（）或cudaMemset（）一起使用时，系统可能会选择从主机或设备访问源或目标，这会限制并发CPU访问该数据，而cudaMemcpy（） 或cudaMemset（）正在执行。 有关更多详细信息，请参阅使用托管内存的Memcpy（）/ Memset（）行为。
当GPU对于concurrentManagedAccess属性设置为0的设备处于活动状态时，不允许CPU访问任何托管分配或变量。在这些系统上，并发CPU / GPU访问（即使对于不同的托管内存分配）也会导致分段错误，因为 该页面被认为是CPU无法访问的页面。
__device__ __managed__ int x, y = 2;
__global__ void kernel() {
x = 10;
}
int main() {
kernel << < 1, 1 >> >();
y = 20; // Error on GPUs not supporting concurrent access
cudaDeviceSynchronize();
return 0;
}
在上面的例子中，GPU程序内核在CPU触摸y时仍处于活动状态。 （请注意它在cudaDeviceSynchronize（）之前是如何发生的。）由于GPU页面错误功能，代码在计算能力6.x的设备上成功运行，这提升了对同时访问的所有限制。 但是，即使CPU正在访问与GPU不同的数据，这种内存访问在6.x之前的体系结构中也是无效的。 在访问y之前，程序必须显式地与GPU同步：
__device__ __managed__ int x, y = 2;
__global__ void kernel() {
x = 10;
}
int main() {
kernel << < 1, 1 >> >();
cudaDeviceSynchronize();
y = 20; // Success on GPUs not supporing concurrent access
return 0;
}
如本例所示，在具有6.x GPU前体系结构的系统上，不管GPU内核是否实际触及相同的数据，CPU线程都可能无法访问执行内核启动和后续同步调用之间的任何托管数据（或 任何管理数据）。 并发CPU和GPU访问的可能性足以满足进程级异常的发生。