【NCCL】Merged Device(合并设备)和bond的区别
在现代高性能计算环境中,节点可能有多个网络接口卡(NIC)或多个设备用于网络通信。NCCL 支持将这些设备合并为一个逻辑连接,以提高带宽和性能。因此,通信的两端都需要知道对方有多少个设备,这样才能正确地建立连接和分配资源。。
NCCL 中将多个设备合并为一个逻辑连接,这种合并被称为"Merged Device"(合并设备)。
"Merged Device"和bond 的区别
"Merged Device"与网络绑定(bonding)之间的异同:
相似之处
- 聚合带宽: 两者都旨在通过组合多个物理网络接口来增加总带宽。
- 负载分担: 都可以在多个物理链路之间分配流量以提高整体吞吐量。
- 透明性: 对于上层应用来说,它们都表现为单个逻辑接口。
主要区别
1. 实现层面
- Bonding(网卡绑定): 是操作系统或网络硬件层面的功能,工作在网络协议栈的较低层次。它将多个物理网卡绑定成一个逻辑接口,对于IP层及以上是透明的。
- NCCL Merged Device: 是NCCL库内部实现的一种机制,专门针对GPU通信优化,在应用程序层面进行管理和调度。
2. 控制粒度
- Bonding: 通常在整个网络堆栈上应用统一策略,比如轮询、主备等模式。
- NCCL Merged Device: 提供更细粒度的控制,可以根据不同的通信模式和GPU拓扑结构动态选择最佳路径。
3. 协议支持
- Bonding: 工作在以太网级别,适用于所有基于IP的通信。
- NCCL Merged Device: 专门为RDMA/InfiniBand优化设计,能够利用RDMA的零拷贝特性和远程直接内存访问能力。
4. 故障恢复
- Bonding: 通常有标准的故障检测和恢复机制。
- NCCL Merged Device: 具有针对GPU集群环境定制的故障处理逻辑,可以更好地适应HPC环境的需求。
5. 配置方式
- Bonding: 通过系统级配置(如/etc/network/interfaces或NetworkManager)进行设置。
- NCCL Merged Device: 由NCCL自动探测和配置,或者通过NCCL特定的环境变量调整。
NCCL Merged Device 的特殊优势
- RDMA优化: 可以为每个设备单独建立RDMA连接,然后智能地在这些连接之间分配通信任务。
- NUMA感知: 能够结合CPU NUMA拓扑,将通信任务分配到最接近GPU的网络接口上。
- 动态适应: 可以根据通信模式(如allreduce、allgather等)动态调整数据分布策略。
总的来说,虽然两者都有聚合多个网络接口的目的,但NCCL Merged Device是一种更高层次、更专业化的解决方案,专为GPU集群的高性能计算场景而设计。
NCCL Merged Device实现
在net_ib.cc文件中实现,主要通过以下结构:
struct ncclIbMergedDev {int ndevs; // 包含的物理设备数量int devs[NCCL_IB_MAX_DEVS_PER_NIC]; // 物理设备索引数组char devName[MAXNAMESIZE]; // 合并设备名称int speed; // 总带宽速度
};当NCCL初始化时,它会检测系统中的所有InfiniBand设备,并根据设备名称和PCI路径判断哪些设备属于同一个NIC的不同端口。如果启用了NCCL_IB_MERGE_NICS参数(默认启用),NCCL会将这些设备合并为一个Merged Device。
合并过程的关键代码如下:
// 将多个物理设备合并为一个逻辑设备
if (mergedDev == ncclNMergedIbDevs) {// 没有找到匹配的设备,创建新的mergedDev条目ncclIbMergedDevs[mergedDev].ndevs = 1;ncclIbMergedDevs[mergedDev].devs[0] = ncclNIbDevs;ncclNMergedIbDevs++;strncpy(ncclIbMergedDevs[mergedDev].devName, ncclIbDevs[ncclNIbDevs].devName, MAXNAMESIZE);
} else {// 找到匹配的设备,添加到现有mergedDev中int ndevs = ncclIbMergedDevs[mergedDev].ndevs;ncclIbMergedDevs[mergedDev].devs[ndevs] = ncclNIbDevs;ncclIbMergedDevs[mergedDev].ndevs++;snprintf(ncclIbMergedDevs[mergedDev].devName + strlen(ncclIbMergedDevs[mergedDev].devName), MAXNAMESIZE+1, "+%s", ncclIbDevs[ncclNIbDevs].devName);
}QP选择机制(数据发送)
在发送数据时,NCCL使用轮询方式在多个QP(Queue Pair)之间分配数据。具体实现在ncclIbMultiSend函数中:
// 根据配置决定是按QP还是按设备分割数据
const int align = 128;
int nqps = ncclParamIbSplitDataOnQps() ? comm->base.nqps : comm->base.ndevs;
for (int i = 0; i < nqps; i++) {int qpIndex = comm->base.qpIndex;ncclIbQp* qp = comm->base.qps + qpIndex;int devIndex = qp->devIndex;// ... 数据发送逻辑 ...// 选择下一个qpIndexcomm->base.qpIndex = (comm->base.qpIndex+1) % comm->base.nqps;
}选择过程的关键点:
- 轮询机制:使用
comm->base.qpIndex跟踪当前使用的QP索引,每次发送后递增并取模 - 数据分割:将要发送的数据按QP数量进行分割,每个QP发送一部分数据
- 对齐处理:确保数据块大小是128字节的倍数,以兼容LL和LL128协议
- 配置选项:通过
NCCL_IB_SPLIT_DATA_ON_QPS参数控制是按QP数量还是设备数量分割数据
这种设计使得NCCL能够:
- 充分利用多端口NIC的所有带宽
- 在多个物理链路间负载均衡
- 提供故障冗余(如果一个QP失败)
- 通过并行发送提高整体吞吐量
与传统的网络绑定(bonding)相比,NCCL Merged Device在应用层实现,能够更好地适应GPU通信的特点,并提供更精细的控制。
