依据物理拓扑结构,来进行通信,可达到性能更优,这也是Nvidia NCCL的核心功能。
视频教程
2.1 NCCL拓扑识别感知总览 源码解读_哔哩哔哩_bilibili
NCCL拓扑识别的整体思路:
1、物理拓扑构建
2、通信路径计算(每个GPU/网卡到其它GPU,网卡的最优路径。)
3、逻辑拓扑构建(通信通道检索)
哈哈哈,整体思路是不是特别简单。就是先获取物理拓扑图,然后计算通信路径(方便逻辑拓扑构建),根据通信路径构建逻辑拓扑,例如ring,tree逻辑拓扑,指明哪个GPU和哪个GPU通信。
具体每个细节下次分享会讲,这次大家先留个映象。
源码速递
源码位置:nccl-master\src\init.cc
1、物理拓扑构建:ncclTopoGetSystem()
2、通信路径计算:ncclTopoComputePaths()
3、逻辑拓扑构建(通信通道检索):ncclTopoCompute()
static ncclResult_t initTransportsRank(struct ncclComm* comm, struct ncclComm* parent = NULL) {
// 其它代码
// 获取系统的拓扑信息,并存储在comm的topo成员中
NCCLCHECKGOTO(ncclTopoGetSystem(comm, &comm->topo), ret, fail);
// 在已获取的拓扑中,计算GPU和NIC之间的路径
NCCLCHECKGOTO(ncclTopoComputePaths(comm->topo, comm), ret, fail);
// 根据计算结果,移除不可访问的GPU和未使用的NIC
NCCLCHECKGOTO(ncclTopoTrimSystem(comm->topo, comm), ret, fail);
// 在移除不可访问的组件后,重新计算路径
NCCLCHECKGOTO(ncclTopoComputePaths(comm->topo, comm), ret, fail);
// 初始化拓扑搜索
NCCLCHECKGOTO(ncclTopoSearchInit(comm->topo), ret, fail);
// 打印最终的拓扑结构,用于调试或信息展示
NCCLCHECKGOTO(ncclTopoPrint(comm->topo), ret, fail);
// 获取与当前GPU本地化的CPU亲和性,即哪些CPU与当前GPU通信效率最高
NCCLCHECKGOTO(ncclTopoGetCpuAffinity(comm->topo, comm->rank, &comm->cpuAffinity), ret, fail);
// 如果找到了与GPU匹配的CPU亲和性(即找到了可用的CPU集合)
if (CPU_COUNT(&comm->cpuAffinity)) {
// 保存当前线程的CPU亲和性设置(可能是为了之后恢复)
sched_getaffinity(0, sizeof(cpu_set_t), &affinitySave);
// 将当前线程的CPU亲和性设置为与GPU匹配的CPU集合
sched_setaffinity(0, sizeof(cpu_set_t), &comm->cpuAffinity);
}
// 检查本地是否支持CollNet(NCCL的一种优化)
if (collNetSupport(comm)) {
// 获取环境变量NCCL_COLLNET_ENABLE的值,决定是否启用CollNet
const char *collNetEnable = ncclGetEnv("NCCL_COLLNET_ENABLE");
if (collNetEnable != NULL) {
// 如果环境变量已设置,打印信息到日志或控制台
INFO(NCCL_ALL, "NCCL_COLLNET_ENABLE set by environment to %s.", collNetEnable);
// 如果环境变量值为"1",则启用CollNet支持
if (strcmp(collNetEnable, "1") == 0) {
comm->collNetSupport = 1;
}
}
}
// 初始化Nvls支持第三代NVSwitch系统(NVLink4)
NCCLCHECK(ncclNvlsInit(comm));
// 初始化环图结构,用于表示环形的通信模式
memset(&ringGraph, 0, sizeof(struct ncclTopoGraph));
ringGraph.id = 0;
ringGraph.pattern = NCCL_TOPO_PATTERN_RING;
ringGraph.minChannels = 1;
ringGraph.maxChannels = MAXCHANNELS/2;
// 在已获取的拓扑中,计算环图的通信信息
NCCLCHECKGOTO(ncclTopoCompute(comm->topo, &ringGraph), ret, fail);
// 打印环图的拓扑结构,用于调试或信息展示
NCCLCHECKGOTO(ncclTopoPrintGraph(comm->topo, &ringGraph), ret, fail);
// 其它代码
}
下次分享:
物理拓扑构建