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 //Static
 3 __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

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

     特点:具有纹理缓存,只读。

TNE END

转载于:https://www.cnblogs.com/traceorigin/archive/2013/04/11/3015482.html

### 回答1: CUDA共享内存是一种特殊的内存类型,它可以在同一个线程块内的线程之间共享数据。这种内存类型的访问速度非常快,因为它是在GPU芯片上的SRAM中实现的。使用共享内存可以有效地减少全局内存的访问,从而提高CUDA程序的性能。共享内存的大小是有限制的,通常为每个线程块的总共享内存大小的一半。因此,在使用共享内存时需要仔细考虑内存的使用情况,以避免内存溢出和性能下降。 ### 回答2: CUDA shared memory是一种专门用于加速GPU并行计算的高速缓存区域。它位于GPU的多个处理核心之间共享,并在同一个线程块中的线程之间交流数据。相比于全局内存,shared memory具有更低的访问延迟和更高的带宽。 shared memory可以通过声明__shared__关键字来定义,并通过静态分配的方式进行初始化。每个线程块都具有自己独立的shared memory空间,其大小在编译时确定,但最大限制为48KB。 shared memory的主要优点是其高带宽和低延迟。由于其位于多个处理核心之间共享,可以实现线程之间的快速数据交换。通过将计算中频繁使用的数据存储在shared memory中,可以减少从全局内存中读取数据所需的时间。这对于那些具有访存限制的算法,如矩阵乘法和图像处理等,非常有用。 使用shared memory还可以避免线程间的数据冗余读取,从而提高整体的并行计算效率。当多个线程需要访问相同的数据时,可以将这些数据存储在shared memory中,以便线程之间进行共享,从而减少了重复的全局内存访问。 但shared memory也有一些限制和需要注意的地方。首先,shared memory的大小是有限的,需要根据具体的算法和硬件限制进行适当调整。其次,由于其共享的特性,需要确保线程之间的数据同步。最后,使用shared memory时需要注意避免bank conflict,即多个线程同时访问同一个shared memory bank造成的资源竞争,从而导致性能下降。 综上所述,CUDA shared memory在GPU并行计算中具有重要的作用。通过使用shared memory,可以有效减少全局内存访问、提高数据交换速度和并行计算效率,从而加速GPU上的并行计算任务。 ### 回答3: CUDA共享内存(shared memory)是指在CUDA程序中使用的一种特殊的内存空间。它是GPU上的一块高速、低延迟的内存,被用来在同一个线程块(thread block)中的线程之间进行数据共享。 与全局内存相比,共享内存的访问速度更快,读写延迟更低。这是因为共享内存位于SM(Streaming Multiprocessor)内部,可以直接被SM访问,而全局内存则需要通过PCIe总线与主机内存进行通信。 使用共享内存可以提高应用程序性能的原因之一是避免了全局内存的频繁访问。当多个线程需要读写同一个数据时,如果每个线程都从全局内存中读取/写入,会导致内存带宽饱和,限制了整体性能。而将这些数据缓存在共享内存中,可以减少对全局内存的访问次数,提高内存带宽的利用率。 除此之外,共享内存的另一个重要特性是可以用作线程间的通信机制。在同一个线程块中的线程可以通过共享内存交换数据,而无需利用全局内存作为中介。这使得线程之间的协作变得更加高效和灵活。 然而,共享内存也有一些限制。首先,共享内存的大小是有限的,通常为每个SM的一定容量(如16KB或48KB)。其次,共享内存的生命周期与线程块相同,每个线程块结束后,共享内存中的数据将被销毁。 在编写CUDA程序时,可以使用__shared__关键字来声明共享内存。同时需要注意,合理地使用共享内存,并避免冲突和竞争条件,才能充分发挥共享内存的优势,提高CUDA程序的性能。
评论
添加红包

请填写红包祝福语或标题

红包个数最小为10个

红包金额最低5元

当前余额3.43前往充值 >
需支付:10.00
成就一亿技术人!
领取后你会自动成为博主和红包主的粉丝 规则
hope_wisdom
发出的红包
实付
使用余额支付
点击重新获取
扫码支付
钱包余额 0

抵扣说明:

1.余额是钱包充值的虚拟货币,按照1:1的比例进行支付金额的抵扣。
2.余额无法直接购买下载,可以购买VIP、付费专栏及课程。

余额充值