CUDA memory

原文链接

CUDA存储器类型：

每个线程拥有自己的register and loacal memory;

每个线程块拥有一块shared memory;

所有线程都可以访问global memory;

还有，可以被所有线程访问的只读存储器：constant memory and texture memory

1、寄存器Register

　　寄存器是GPU上的高速缓存器，其基本单元是寄存器文件，每个寄存器文件大小为32bit.

　　Kernel中的局部(简单类型)变量第一选择是被分配到Register中。

　　特点：每个线程私有，速度快。

2、局部存储器 local memory

　　当register耗尽时，数据将被存储到local memory。如果每个线程中使用了过多的寄存器，或声明了大型结构体或数组，或编译器无法确定数组大小，线程的私有数据就会被分配到local 　　memory中。

　　特点：每个线程私有；没有缓存，慢。

　　注：在声明局部变量时，尽量使变量可以分配到register。如：

　　unsigned int mt[3];

　　改为：　unsigned int mt0, mt1, mt2;

3、共享存储器 shared memory

　　可以被同一block中的所有线程读写

　　特点：block中的线程共有；访问共享存储器几乎与register一样快.

 1 //u(i)= u(i)^2 + u(i-1)2 //Static3 __global__ example(float* u) {4     int i=threadIdx.x;5     __shared__ int tmp[4];6      tmp[i]=u[i];7      u[i]=tmp[i]*tmp[i]+tmp[3-i];8 }9
10 int main() {
11     float hostU[4] = {1, 2, 3, 4};
12      float* devU;
13     size_t size = sizeof(float)*4;
14     cudaMalloc(&devU, size);
15     cudaMemcpy(devU, hostU, size,
16     cudaMemcpyHostToDevice);
17      example<<<1,4>>>(devU, devV);
18     cudaMemcpy(hostU, devU, size,
19     cudaMemcpyDeviceToHost);
20     cudaFree(devU);
21     return 0;
22 }
23
24 //Dynamic
25 extern __shared__ int tmp[];
26
27 __global__ example(float* u) {
28     int i=threadIdx.x;
29      tmp[i]=u[i];
30      u[i]=tmp[i]*tmp[i]+tmp[3-i];
31 }
32
33 int main() {
34     float hostU[4] = {1, 2, 3, 4};
35     float* devU;
36     size_t size = sizeof(float)*4;
37     cudaMalloc(&devU, size);
38     cudaMemcpy(devU, hostU, size, cudaMemcpyHostToDevice);
39      example<<<1,4,size>>>(devU, devV);
40     cudaMemcpy(hostU, devU, size, cudaMemcpyDeviceToHost);
41     cudaFree(devU);
42     return 0;
43 }

4、全局存储器 global memory

　　特点：所有线程都可以访问；没有缓存

//Dynamic
__global__ add4f(float* u, float* v) {
int i=threadIdx.x;u[i]+=v[i];
}
int main() {float hostU[4] = {1, 2, 3, 4};float hostV[4] = {1, 2, 3, 4};float* devU, devV;size_t size = sizeof(float)*4;cudaMalloc(&devU, size);cudaMalloc(&devV, size);cudaMemcpy(devU, hostU, size,cudaMemcpyHostToDevice);cudaMemcpy(devV, hostV, size,cudaMemcpyHostToDevice);add4f<<<1,4>>>(devU, devV);cudaMemcpy(hostU, devU, size,cudaMemcpyDeviceToHost);cudaFree(devV);cudaFree(devU);return 0;
}//static
__device__ float devU[4];
__device__ float devV[4];__global__ addUV() {
int i=threadIdx.x;devU[i]+=devV[i];
}int main() {float hostU[4] = {1, 2, 3, 4};float hostV[4] = {1, 2, 3, 4};size_t size = sizeof(float)*4;cudaMemcpyToSymbol(devU, hostU, size, 0, cudaMemcpyHostToDevice);cudaMemcpyToSymbol(devV, hostV, size, 0, cudaMemcpyHostToDevice);addUV<<<1,4>>>();cudaMemcpyFromSymbol(hostU, devU, size, 0, cudaMemcpyDeviceToHost);return 0;
}

5、常数存储器constant memory

　　用于存储访问频繁的只读参数

　　特点：只读；有缓存；空间小(64KB)

　　注：定义常数存储器时，需要将其定义在所有函数之外，作用于整个文件

1 __constant__ int devVar;
2 cudaMemcpyToSymbol(devVar, hostVar, sizeof(int), 0, cudaMemcpyHostToDevice)
3 cudaMemcpyFromSymbol(hostVar, devVar, sizeof(int), 0, cudaMemcpyDeviceToHost)

6、纹理存储器 texture memory

是一种只读存储器，其中的数据以一维、二维或者三维数组的形式存储在显存中。在通用计算中，其适合实现图像处理和查找，对大量数据的随机访问和非对齐访问也有良好的加速效果。

特点：具有纹理缓存，只读。

转载于:https://www.cnblogs.com/liangliangdetianxia/p/4198850.html

CUDA memory相关推荐

cuda Memory Fence Functions
thread对内存读写的安全 void __threadfence_block(); void __threadfence(); void __threadfence_system(); 参考: ht ...
CUDA系列学习（四）Parallel Task类型与 Memory Allocation
本文为CUDA系列学习第四讲,首先介绍了Parallel communication patterns的几种形式(map, gather, scatter, stencil, transpose), ...
cuda的global memory介绍
CUDA Memory Model 对于程序员来说,memory可以分为下面两类: Programmable:我们可以灵活操作的部分. Non-programmable:不能操作,由一套自动机制来达到 ...
【解决汇总：待更新】CUDA error: an illegal memory access was encountered with channels_last
为方便查看,就直接贴图和网址了.仅供学习使用,如有侵权,请联系我删除. 一.调试方法: 链接原文: Recently, I encountered "an illegal memory a ...
CUDA(六). 从并行排序方法理解并行化思维——冒泡、归并、双调排序的GPU实现
在第五讲中我们学习了GPU三个重要的基础并行算法: Reduce, Scan 和 Histogram,分析了其作用与串并行实现方法. 在第六讲中,本文以冒泡排序 Bubble Sort.归并排序 M ...
CUDA系列学习（三）GPU设计与结构QA coding练习
啥?你把CUDA系列学习(一),(二)都看完了还不知道為什麼要用GPU提速? 是啊..经微博上的反馈我默默感觉到提出这样问题的小伙伴不在少数,但是更多小伙伴应该是看了(一)就感觉离自己太远所以赶紧撤粉 ...
CUDA 编程简介（下）
文章目录 Memory shared memory global memory Transfer Data 异步预取 Threads thread block warp GPU 性能查看性能测试性 ...
CPU与GPU统一虚拟内存(CUDA UM)原理
CPU与GPU的统一内存(CUDA Unified Memory)原理文章目录 CPU与GPU的统一内存(CUDA Unified Memory)原理一.UM下的CUDA编程二.UM的实现原理 ...
CUDA学习(四):CUDA编程七个步骤
博主CUDA学习系列汇总传送门(持续更新):编程语言|CUDA入门文章目录一.cudaMalloc.cudaMemcpy和cudaFree 介绍二.CUDA编程七步曲本章节学习内容: 1.CU ...
linux .sh文件命令如何写,Linux下面使用命令如何运行.sh文件的两种解决办法
SpringMVC 参数注入写一个web service, 总是400. 说是request有问题,server不识别.然而检查了很多次都没问题.最终问题指向spring对参数的解析和注入. 一个c ...

CUDA memory

CUDA memory相关推荐

最新文章

热门文章