CUDA 6.0统一寻址
来源:互联网 发布:剑网3军娘捏脸数据 编辑:程序博客网 时间:2024/05/21 04:19
大家知道,Nvidia公司发布的CUDA6开发包拥有一个新特性,就是“统一内存寻址”,那究竟统一内存寻址有什么特殊的地方呢?我们编写CUDA代码跟以前有什么区别呢?现在拥有的GPU架构能够很好的支持吗?我们带着这些问题开始我们的话题。
从名字上看,统一内存寻址就是将CPU端的内存同GPU显存统一起来,使得程序猿在编写代码的时候不用明显的使用诸如cudaMalloc或者cudaMemcpy等操作显存的指令,而能够在Kernel函数中直接使用定义的变量。
例子如下:
原始代码(CUDA6.0文档)
统一寻址的代码:
从上面不同的代码可以看出,统一寻址后的代码更简洁,使用了函数cudaMallocManaged()开辟一块存储空间,无论是在Kernel函数中还是main函数中,都可以使用这块内存,达到了统一寻址的目的。
这里有一个需要注意的地方就是main函数在调用Kernel函数之后,使用了一个同步函数。仔细思考后就会有所领悟——既然这块存储空间既可以被Kernel函数访问,也可以被main函数访问,为了解决访问冲突的问题,因此使用了同步函数,使得在Kernel改变变量的值后,main函数才能使用该变量。
上面只是简单的举了个例子,但已经把统一寻址的问题描述的很清楚:
1. 开辟空间所使用的新函数;
2. 变量访问同步。
在CUDA6.0的文档中更加仔细的描述了相关的问题,包括流和多GPU条件下使用统一寻址的问题等,其主要针对的还是变量访问同步,大家有兴趣可以参看文档。
同AMD的APU不同,即使是Kepler架构的GPU,也不是真正意义上从硬件手段实现“统一内存寻址”,只是从软件上解决的一种方式,Nvidia公司的新一代GPU产品——Maxwell架构,才真正是从硬件层面实现这一特性。而速度,还有待测试与研究~~
既然“统一内存寻址”是GPU产品所前进的一个方向,势必会改变我们的编程方式,因此,这一话题还是有重大的意义的,希望大家热烈讨论,在讨论中真正掌握这一技术!
从名字上看,统一内存寻址就是将CPU端的内存同GPU显存统一起来,使得程序猿在编写代码的时候不用明显的使用诸如cudaMalloc或者cudaMemcpy等操作显存的指令,而能够在Kernel函数中直接使用定义的变量。
例子如下:
原始代码(CUDA6.0文档)
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
20
21
22
__global__
void
AplusB(
int
*ret,
int
a,
int
b)
{
ret[threadIdx.x] = a + b + threadIdx.x;
}
int
main()
{
int
*ret;
//**************************************
cudaMalloc(&ret, 1000 *
sizeof
(
int
));
AplusB<<<1, 1000>>>(ret, 10, 100);
//**************************************
int
*host_ret = (
int
*)
malloc
(1000 *
sizeof
(
int
));
cudaMemcpy(host_ret, ret, 1000 *
sizeof
(
int
), cudaMemcpyDefault);
for
(
int
i = 0; i < 1000; i++)
printf
(
"%d: A + B = %d\n"
, i, host_ret[i]);
free
(host_ret);
cudaFree(ret);
return
0;
}
统一寻址的代码:
1
2
3
4
5
6
7
8
9
10
11
12
13
14
15
16
17
18
19
__global__
void
AplusB(
int
*ret,
int
a,
int
b)
{
ret[threadIdx.x] = a + b + threadIdx.x;
}
int
main()
{
int
*ret;
//***********************************************
cudaMallocManaged(&ret, 1000 *
sizeof
(
int
));
AplusB<<<1, 1000>>>(ret, 10, 100);
//***********************************************
cudaDeviceSynchronize();
for
(
int
i = 0; i < 1000; i++)
printf
(
"%d: A + B = %d\n"
, i, ret[i]);
cudaFree(ret);
return
0;
}
从上面不同的代码可以看出,统一寻址后的代码更简洁,使用了函数cudaMallocManaged()开辟一块存储空间,无论是在Kernel函数中还是main函数中,都可以使用这块内存,达到了统一寻址的目的。
这里有一个需要注意的地方就是main函数在调用Kernel函数之后,使用了一个同步函数。仔细思考后就会有所领悟——既然这块存储空间既可以被Kernel函数访问,也可以被main函数访问,为了解决访问冲突的问题,因此使用了同步函数,使得在Kernel改变变量的值后,main函数才能使用该变量。
上面只是简单的举了个例子,但已经把统一寻址的问题描述的很清楚:
1. 开辟空间所使用的新函数;
2. 变量访问同步。
在CUDA6.0的文档中更加仔细的描述了相关的问题,包括流和多GPU条件下使用统一寻址的问题等,其主要针对的还是变量访问同步,大家有兴趣可以参看文档。
同AMD的APU不同,即使是Kepler架构的GPU,也不是真正意义上从硬件手段实现“统一内存寻址”,只是从软件上解决的一种方式,Nvidia公司的新一代GPU产品——Maxwell架构,才真正是从硬件层面实现这一特性。而速度,还有待测试与研究~~
既然“统一内存寻址”是GPU产品所前进的一个方向,势必会改变我们的编程方式,因此,这一话题还是有重大的意义的,希望大家热烈讨论,在讨论中真正掌握这一技术!
0 0
- CUDA 6.0统一寻址
- CUDA 6.0 统一寻址
- CUDA 6.0 统一内存寻址
- NVIDIA正式宣布CUDA 6.0:支持统一寻址!
- NVIDIA正式宣布CUDA 6.0:支持统一寻址!
- NVIDIA正式宣布CUDA 6.0:支持统一寻址!
- 统一寻址
- CUDA Unified Memory统一内存使用注意
- CUDA Unified Memory统一内存使用注意
- NVIDIA CUDA统一计算设备架构编程手册(1)
- CUDA统一内存、零复制内存、锁页内存
- 寻址
- 寻址
- IO端口与寻址(统一编址与独立编址)
- CUDA
- CUDA
- CUDA
- CUDA
- 大数据时代基本观点
- Qt 5.3更新无数,更改C++控制台输出最为赞
- 2014 BNU邀请赛F题(枚举)
- MIME Types类型
- 如何利用宝贝上下架时间获取流量
- CUDA 6.0统一寻址
- Python 输出用于 Paraview 后处理的 vtk 文件
- LeetCode: Maximum Subarray [052]
- 学习DTD
- 深度网络学习软件资源
- 分形问题
- CUDA学习之浅谈cuBLAS
- 数据类型中 char,nchar,nvarchar,varchar 的区别
- unity character controller 碰撞