一、parallel communication patterns   并行通信模式

Map:映射,在特定的位置读取和写入。

Gather:收集,从多个不同的位置读入,写入一个位置。

Scatter:分发,写入多个位置。

Transpose转置

结构数组缩写为AOS,数组结构缩写为SOA

转置运算是指任务重新排序内存中的数据元素

Intro to Parallel Programming CUDA-第二单元

Stencil模板

能够在输出结果为每一个元素生成一个结果

Intro to Parallel Programming CUDA-第二单元

Intro to Parallel Programming CUDA-第二单元

二、流处理器Stream Multiprocessors

所有在同一个线程块中的线程运行在同一个SM上

在一个内核中的所有块,在下个内核中任意块启动前都已完成

三、共享内存

共享内存是在一个块的线程之间共享

CUDA中的线程协作主要是通过共享内存实现的。使用关键字“__share__”声明共享变量,将使这个变量驻留在共享内存中,该变量具有以下特征:

  • 位于线程块的共享存储器空间中
  • 与线程块具有相同的生命周期
  • 仅可通过块内的所有线程访问

对于GPU上启动的每个线程块,CUDA C编译器都将创建该变量的一个副本。 线程块中的每个线程都共享这块内存,但线程却无法看到也不能修改其他线程块的变量副本。 这就使得一个线程块中的多个线程能够在计算上进行通信和协作。而且,共享内存缓冲区驻留在物理GPU上,在访问共享内存时的延迟要远远低于访问普通缓冲区的延迟,使得共享内存的访问非常高效。

线程同步机制“__syncthreads()”

关键字“__share__”只是声明了共享变量,位于同一个线程块中的不同线程都可以访问该变量,如果没有同步机制,将会发生竞态条件 (Race Condition),导致错误的运行结果。

CUDA确保同步的方法是调用“__syncthreads()”。__syncthreads()将确保线程块中的每个线程都执行完 __syncthreads()前面的语句后,才会执行下一条语句。

一个线程可以访问局部内存、一个线程块的共享内存和所有线程都可以访问的全局内存

速度方面:local>shared>global>cpu

Intro to Parallel Programming CUDA-第二单元

四、屏障

Intro to Parallel Programming CUDA-第二单元

__syncthreads()的线程需要其他可以到达该点的线程,而不是等待块内所有其他线程

一个高端的GPU每秒可以进行超过3万亿次的数学计算

五、Global Memory

如果想操作全局变量必须传递一个指向该内存的指针,如:

Intro to Parallel Programming CUDA-第二单元

诀窍:只可以在内核中传递局部变量,然后传入一个指针,要在内核外分配和初始化全局变量

相关文章:

  • 2021-09-28
  • 2021-10-23
  • 2022-12-23
  • 2021-12-10
  • 2021-11-28
  • 2022-12-23
  • 2021-08-14
  • 2021-06-15
猜你喜欢
  • 2021-04-20
  • 2022-12-23
  • 2022-12-23
  • 2021-08-13
  • 2021-11-12
  • 2021-06-02
  • 2021-04-10
相关资源
相似解决方案