【问题标题】:How do GPU cores communicate with each other?GPU 内核如何相互通信?
【发布时间】:2020-11-15 16:07:24
【问题描述】:

GPU 在用于通用计算时,非常重视 SIMD 和 SIMT 的细粒度并行性。它们在具有高算术强度的常规数字处理工作负载上表现最佳。

尽管如此,要适用于尽可能多的工作负载,它们还必须能够实现粗粒度 MIMD 并行性,其中不同的内核在不同的数据块上执行不同的指令流。

这意味着 GPU 上的不同核心在执行不同的指令流后必须相互同步。他们是怎么做到的?

在 CPU 上,答案将是缓存一致性以及一组选择与 CAS 或 LL/SC 等通信原语良好配合的通信原语。但据我了解,GPU 没有缓存一致性——避免这种开销是它们首先比 CPU 更高效的最大原因。

那么 GPU 内核使用什么方法来相互同步呢?如果他们如何交换数据的答案是写入共享主内存,那么他们如何同步以便发送者可以通知接收者何时读取数据?

如果答案取决于特定的架构,那么我对支持 CUDA 的现代 Nvidia GPU 特别感兴趣。

编辑:来自 Booo 链接的文档,这是我目前的理解:

他们似乎使用“流”这个词来表示大量同步完成的事情(包括像 SIMD 这样的细粒度并行性);那么问题是如何在多个流之间同步/通信。

正如我推测的那样,这比在 CPU 上要明确得多。特别是,他们谈论:

  • 页面锁定内存
  • cudaDeviceSynchronize ()
  • cudaStreamSynchronize (streamid)
  • cudaEventSynchronize(事件)

因此,流可以通过将数据写入主内存(或 L3 缓存?)进行通信,而 CPU 上没有缓存一致性,取而代之的是锁定内存页面和/或显式同步 API。

【问题讨论】:

  • 当您提到 GPU 上的“核心”时。您是指 Nvidia GPU 上的流式多处理器 (SM) 吗?还是计算算术流水线的“CUDA 核心”?我认为 CUDA 无法让您控制 CUDA 核心级别的任何并行性。
  • 流之间如何同步? some slides 提供了简短的答案:您可以使用事件进行同步,或者 w.r.t.一个特定的流,或整个设备。
  • @Booo 我不太了解 GPU 使用的所有术语,但我的意思是在最粗略的粒度上,即 MIMD 而不是 SIMD 或 SIMT 的级别。感谢您的链接!看起来不错,将总结发现。

标签: multithreading synchronization nvidia gpgpu


【解决方案1】:

我的理解是有几种方法可以使用 CUDA“同步”:

  • CUDA 流(在功能级别):cudaDeviceSynchronize() 在整个设备上同步。此外,您可以将特定流与cudaStreamSynchronize(cudaStream_t stream) 同步,或将嵌入在某些流中的事件与cudaEventSynchronize(cudaEvent_t event) 同步。 Ref 1, Ref 2.

  • 协作组(>CUDA 9.0 和 >CC 3.0):您可以在组级别进行同步,组可以是一组合并的线程、一个线程块或跨越多个设备的网格。这要灵活得多。使用

    定义您自己的组

    (1) auto group = cooperative_groups::coalesced_threads() 用于当前合并的线程集,或

    (2)auto group = cooperative_groups::this_thread_block()对于当前线程块,可以在块内进一步定义细粒度组如auto group_warp = cooperative_groups::tiled_partition<32>(group),或者

    (3) auto group = cooperative_groups::this_grid()auto group = cooperative_groups::this_multi_grid() 用于跨多个设备的网格。

    然后,您可以致电group.sync() 进行同步。您需要有一个支持cooperativeLaunchcooperativeMultiDeviceLaunch 的设备。请注意,对于协作组,您已经可以使用__syncthreads() 在共享内存中执行传统的块级同步。 Ref 1, Ref 2.

【讨论】:

    猜你喜欢
    • 2012-03-22
    • 2011-06-21
    • 1970-01-01
    • 2023-03-17
    • 1970-01-01
    • 1970-01-01
    • 2010-10-12
    • 1970-01-01
    • 2012-04-21
    相关资源
    最近更新 更多