【微知】NVIDIA GPU间通信库NCCL集合通信库的一些基础信息?5种操作?什么是reduce?

背景

NCCL上承CUDA程序,下启RDMA(不一定是rdma)通信,来提供GPU到GPU的集合通信,从而加速GPU间通信。

关键信息

  • NVIDIA Collective Communications Library(NCCL,发音为“Nickel”)是一个库,提供拓扑感知的 GPU 间通信基元
  • 它不是一个成熟的并行编程框架;相反,它是一个专注于加速 GPU 间通信的库。
  • 五种集合通信原语
    • AllReduce 全部减少(规约)
    • Broadcast 广播
    • Reduce 减少
    • AllGather 全聚集
    • ReduceScatter 减少散射
  • 什么是“local reductions” ?在并行计算和分布式系统中,“local reductions” 是指的是在每个计算节点(或设备)内部对数据进行的归约操作。归约操作(reduction)是一种常见的并行计算模式,用于对一组数据执行某种操作(如求和、求最大值、求最小值等),并将结果归约为一个单一的值或一组值。每个设备(例如 GPU)内部独立完成的归约操作,而不是跨设备的全局归约操作。部归约是并行计算中常见的优化手段,可以减少通信开销,提高整体计算效率
  • 集合通信的关键要素:通信处理器之间的紧密同步。通过 CUDA 内存复制和 CUDA 内核组合来实现,以实现局部归约。另一方面,NCCL 在单个内核中实现每个集合,同时处理通信和计算。并听过快速同步,最大限度地减少达到峰值带宽所需的资源。
  • NCCL 在节点内和节点之间为多个 GPU 提供快速集合。它支持各种互连技术,包括 PCIe、NVLINK、InfiniBand Verbs 和 IP 套接字。
  • NCCL 紧跟 MPI(消息传递接口)定义的流行集合 API。因此,任何熟悉 MPI 的人都会发现 NCCL 的 API 使用起来非常自然。
  • NCCL 是一个通信库,为高性能应用程序提供优化的 GPU 到 GPU 通信。它不像 MPI 那样提供包括进程启动器和管理器的并行环境。因此,NCCL 依赖于应用程序的进程管理系统和 CPU 端通信系统来获得自己的引导程序。
  • NCCL 不在 GPU 之间提供安全的网络通信。

几种集合通信源语

Reduce 归约

(这里归约不一定准确 大概表达归一和约的意思,约在小学学习是“压榨掉共同的部分”,比如公约数)

Reduce作执行与 AllReduce 相同的操作(操作是跨设备对数据(例如 sum、min、max)执行归约),但仅将结果存储在指定根排名的接收缓冲区中。
在这里插入图片描述

Reduce 后,在进行 Broadcast 等同于 AllReduce
root 参数是等级之一(不是设备编号),因此会受到不同等级到设备映射的影响。

AllReduce 全归约

AllReduce操作是跨设备对数据(例如 sum、min、max)执行缩减,并将结果存储在每个排名的接收缓冲区中。
在 k 个等级之间的 sum allreduce作中,每个等级将提供一个包含 N 个值的数组,并在 N 个值的数组中接收相同的结果,其中 out[i] = in0[i]+in1[i]+…+in(k-1)[i]
在这里插入图片描述

Broadcast 广播

Broadcast作将 N 元素缓冲区从根 rank 复制到所有 rank。
在这里插入图片描述

AllGather 全聚合

AllGather作将 k 个排名中的 N 个值收集到大小为 k*N 的输出缓冲区中,并将该结果分发到所有rank。输出按 rank 索引排序。因此,AllGather会受到不同等级到设备映射的影响。
在这里插入图片描述

ReduceScatter 归约散射

ReduceScatter作执行与 Reduce 相同的作,不同之处在于结果分散在等级之间大小相等的块中,每个等级根据其等级索引获取一个数据块。

这里有点类似将所有的Reduce之后根据等级拆开 scatter到不同的rank上。
在这里插入图片描述

ReduceScatter作受不同等级到设备映射的影响,因为等级决定了数据布局。

综述:

  • reduce就是把所有的都归约起来,放到某个里面。他是其他操作的根本。allreduce是指归约到某个里面后,然后散射到其他所有gpu。 boardcast不错归约这种算法类的操作,仅仅是“阳光普照”。 ruduce是众人拾柴火焰高。 allreduce是众人拾柴然后阳光普照。
  • AllGather和ReduceScatter强调的是根据排名仅仅取一部分来做操作。和Linux内存的scatter gather类似。都是对整体一个拆散或者聚合成一个整理。操作单元最大是1个单元。 将每个rank的0.x单元放在一起叫allgather,注意是每个。 将某个单元的分成0.x个小单元放入每个rank叫做allscatter
  • reduce是把大家各自的饼放在一起(归)做压缩饼干(约)。allreduce是压缩后再clone给其他(比copy更贴切)。all都用reduce的东西。boardcast是iclone。
  • gather和scatter中gather是每个人拿出一点饼凑成一个饼(gather)然后clone给每个人;scatter是把每个人的饼拆成有序的多个,然后分别给大家。每个GPU只拿自己序号的,然后合在一起。也可以理解是allreduce的一部分。每个人拿对应rank序号的一部分。
  • Reduce当做一种核心操作。allreduce是reduct后然后all。 allgather类似是gather之后给all。 reducescatter是reduce后scatter

参考链接:
nccl官网介绍:https://developer.nvidia.com/nccl
Overview of NCCL

<think>好的,我现在需要帮助用户解决如何使用nccl-tests测试单机内8张GPUNCCL通信性能。首先,我需要回忆一下nccl-tests的基本信息NCCLNVIDIA集合通信库,用于多GPU之间的高效通信。nccl-tests是NVIDIA提供的测试工具集,用来评估不同通信模式下的性能,比如all_reduce、broadcast等。 用户的需求是在单机内测试8张GPU的性能。根据用户提供的引用内容,特别是引用[3]中提到编译nccl-tests项目和API的使用注意事项,我需要先确保用户正确编译了项目。编译步骤通常包括克隆仓、使用make命令,可能需要指定NCCL的路径,如果用户是默认安装的话可能不需要额外参数。 接下来,用户需要运行测试。常用的测试包括all_reduce、broadcast、all_gather等。对于单机8卡,需要指定np参数为8,并可能使用其他参数如gpu id。例如,运行all_reduce测试的命令可能是./build/all_reduce_perf -b 8 -e 256M -f 2 -g 8。这里的-b表示起始数据大小,-e是结束大小,-f是倍增因子,-g是GPU数量。需要确认这些参数是否正确,以及如何调整以获得不同数据量的测试结果。 用户可能还需要注意环境变量的设置,比如NCCL_DEBUG=INFO可以输出调试信息,帮助确认是否所有GPU都被正确识别和使用。此外,引用[3]提到,ncclGroupEnd()返回时操作可能还未执行完毕,所以测试程序内部应该有同步机制确保测量准确性,但用户可能需要了解测试结果的可靠性。 另外,引用[2]提到多机多卡的对比分析,但用户是单机情况,所以不需要涉及网络配置,但可能需要确保所有GPU在单机内正确连接,比如通过PCIe或NVLink,不同的硬件连接会影响带宽结果。用户可能需要对比理论带宽和实测值,以评估通信效率。 最后,生成相关问题时要围绕测试方法、参数调整、结果解读以及性能优化等方面。例如,如何选择合适的数据大小,如何分析测试结果中的带宽和延迟,以及单机多卡通信的优化方法等。</think>### 如何测试单机内8张GPUNCCL通信性能 #### 1. 安装与编译 nccl-tests - **克隆仓**: ```bash git clone https://github.com/NVIDIA/nccl-tests.git cd nccl-tests ``` - **编译项目**(需确保已安装NCCL和CUDA): ```bash make NCCL_HOME=/path/to/nccl CUDA_HOME=/path/to/cuda ``` 若NCCL和CUDA安装在默认路径(如`/usr/local`),可省略路径参数直接运行`make`[^3]。 #### 2. 运行测试 - **测试单机8卡**: ```bash # 示例:测试all_reduce性能 ./build/all_reduce_perf -b 8 -e 256M -f 2 -g 8 ``` - `-b`: 起始数据大小(如8字节) - `-e`: 最大数据大小(如256MB) - `-f`: 每次数据量倍增因子(如2倍) - `-g`: 使用的GPU数量(此处为8) - **其他常用测试项**: - `all_gather_perf`: 测试集合通信中的全收集操作 - `broadcast_perf`: 测试广播性能 - `reduce_scatter_perf`: 测试分散归约性能 #### 3. 关键参数说明 | 参数 | 作用 | 示例值 | |------|------|--------| | `-g` | GPU数量 | `8` | | `-b` | 最小数据量 | `8`(8字节) | | `-e` | 最大数据量 | `256M`(256MB) | | `-c` | 重复测试次数 | `100` | | `-n` | 每次测试的迭代次数 | `20` | #### 4. 结果解读 输出结果包含以下关键指标: - **带宽(Bus Bandwidth)**:单位为GB/s,反映GPU间数据传输效率。 - **算法带宽(Algorithm Bandwidth)**:考虑通信算法的理论优化值。 - **实际带宽(Achieved Bandwidth)**:实测值,接近理论值表明通信效率高。 例如: ``` # nThread 8 nGpus 8 # size count type redop time algbw busbw error 256M 1 float sum 5.432ms 123.4GB/s 115.7GB/s ok ``` #### 5. 注意事项 - **硬件拓扑**:GPU之间的物理连接方式(如NVLink或PCIe)会显著影响带宽[^2]。 - **同步机制**:测试程序会自动同步GPU操作,确保结果准确性。 - **环境变量**: ```bash export NCCL_DEBUG=INFO # 查看详细通信过程 export NCCL_SHM_DISABLE=1 # 禁用共享内存(调试用) ```
评论
添加红包

请填写红包祝福语或标题

红包个数最小为10个

红包金额最低5元

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

抵扣说明:

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

余额充值