CUDA学习(九十六)

简介: 想刷一遍PAT甲级

一致性和并发性:
在计算能力低于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访问的可能性足以满足进程级异常的发生。
timg

相关实践学习
部署Stable Diffusion玩转AI绘画(GPU云服务器)
本实验通过在ECS上从零开始部署Stable Diffusion来进行AI绘画创作,开启AIGC盲盒。
目录
相关文章
|
并行计算 程序员
|
并行计算 API 编译器
|
并行计算 异构计算
|
机器学习/深度学习 缓存 并行计算
|
存储 并行计算
|
并行计算 C语言 编译器
|
存储 并行计算 异构计算
|
并行计算 API
|
并行计算 编译器 存储
|
并行计算 API 编译器
CUDA学习(六十五)
很早之前就发现云栖社区的编辑器有一个Bug,往草稿箱存博客,当草稿箱博客数超过十篇时,无法再选择十篇前的博客进行编辑
2398 0
下一篇
无影云桌面