- 高性能计算
- 通信
- 深度学习
- 分布式训练
【免费下载链接】nccl
Optimized primitives for collective multi-GPU communication
NCCL 提供了一套可配置的诊断日志系统,用于定位初始化挂起、网络通信故障、拓扑探测异常和性能退化等问题,帮助开发者理解 NCCL 在运行时的算法选择、拓扑检测与网络配置决策。本文以 NCCL 用户指南中的 Logging 文档为主线,结合当前仓库(NCCL 2.30.7,见 makefiles/version.mk)中 src/debug.cc 的日志子系统实现,完整讲解五个日志环境变量的用法、日志级别与子系统的底层过滤机制、日志文件写入细节以及时间戳定制方案,并给出五类典型调试场景的推荐配置。
一、日志环境变量总览
NCCL 通过以下环境变量控制日志行为,它们的定义与默认值在 src/debug.cc 中通过DEFINE_NCCL_PARAM声明:
| 变量 | 说明 |
|---|---|
NCCL_DEBUG | 设置日志级别:VERSION、WARN、INFO、TRACE |
NCCL_DEBUG_SUBSYS | 按子系统过滤日志输出(位掩码) |
NCCL_DEBUG_FILE | 将日志写入文件,支持%h(主机名)、%p(进程 PID)占位符 |
NCCL_DEBUG_TIMESTAMP_FORMAT | 设置时间戳格式(strftime 语法) |
NCCL_DEBUG_TIMESTAMP_LEVELS | 选择哪些日志级别带时间戳 |
这些变量除命令行设置外,也可以静态写入/etc/nccl.conf(系统级)或${NCCL_CONF_FILE}指定的配置文件,详见 Environment Variables 文档。
从源码结构看,五条日志链路由以下全局状态驱动(定义于 src/debug.cc):
int ncclDebugLevel = -1; // 当前日志级别 uint64_t ncclDebugMask = 0; // 子系统位掩码 FILE* ncclDebugFile = stdout; // 日志输出目标(默认 stdout) static uint32_t ncclDebugTimestampLevels = 0; // 开启时间戳的级别位图 static char ncclDebugTimestampFormat[256]; // 时间戳格式日志级别和子系统的枚举定义在 src/include/nccl_common.h:
typedef enum { NCCL_LOG_NONE = 0, NCCL_LOG_VERSION = 1, NCCL_LOG_WARN = 2, NCCL_LOG_INFO = 3, NCCL_LOG_ABORT = 4, // 内部级别,未列入用户文档表格 NCCL_LOG_TRACE = 5 } ncclDebugLogLevel; typedef enum { NCCL_INIT = 0x1, NCCL_COLL = 0x2, NCCL_P2P = 0x4, NCCL_SHM = 0x8, NCCL_NET = 0x10, NCCL_GRAPH = 0x20, NCCL_TUNING = 0x40, NCCL_ENV = 0x80, NCCL_ALLOC = 0x100, NCCL_CALL = 0x200, NCCL_PROXY = 0x400, NCCL_NVLS = 0x800, NCCL_BOOTSTRAP = 0x1000, NCCL_REG = 0x2000, NCCL_PROFILE = 0x4000, NCCL_RAS = 0x8000, NCCL_DESTROY = 0x10000, NCCL_ALLOC_HOST = 0x20000, NCCL_ALL = ~0 } ncclDebugLogSubSys;二、日志级别(NCCL_DEBUG)
NCCL 支持多个日志级别,从最简到最详尽依次为:
| 级别 | 说明 | 使用场景 |
|---|---|---|
VERSION | 启动时打印 NCCL 版本 | 验证安装 |
WARN | 警告与错误 | 生产环境最低要求 |
INFO | 详细运行信息 | 诊断运行时问题 |
TRACE | 可重放(replayable)跟踪,含 API 调用记录 | 深度调试 / NCCL 开发 |
设置方式:
NCCL_DEBUG=INFO ./my_app各级别的典型输出如下(摘自日志文档,具体数值、行号和格式可能随 NCCL 版本与环境变化)。
NCCL_DEBUG=VERSION—— 启动时打印版本:
NCCL version 2.30.3+cuda13.0NCCL_DEBUG=WARN—— 打印警告与错误:
[2026-05-05 06:16:47] node-01:189884:189884 [3] plugin/net.cc:334 NCCL WARN Failed to initialize any NET pluginNCCL_DEBUG=INFO—— 打印 NCCL 操作的详细信息:
node-01:3873285:3873285 [0] NCCL INFO Initialized NET plugin IB node-01:3873285:3873285 [0] NCCL INFO Assigned NET plugin IB to comm node-01:3873285:3873285 [0] NCCL INFO Assigned GIN plugin GIN_XXXX to comm node-01:3873285:3873285 [0] NCCL INFO Assigned RMA plugin RMA_XXXX to comm node-01:3873285:3873285 [0] NCCL INFO Using network IB node-01:3873285:3873285 [0] NCCL INFO DMA-BUF is available on GPU device 0NCCL_DEBUG=TRACE NCCL_DEBUG_SUBSYS=CALL—— 函数调用跟踪:
node-01:986207:986207 NCCL CALL ncclGroupStart() node-01:986207:986207 NCCL CALL ncclSend(0,0x...,8388608,7,0,2,0x...,0x...) node-01:986207:986207 NCCL CALL ncclRecv(0,0x...,8388608,7,0,0,0x...,0x...) node-01:986207:986207 NCCL CALL ncclGroupEnd()日志行格式的源码实现
每一行日志的拼装逻辑在 src/debug.cc 的 ncclDebugLogV() 中。以NCCL INFO为例,格式为:
hostname:pid:tid [cudaDev] NCCL INFO <消息>对应代码:
// 追加 hostname、pid、tid(VERSION 级别除外) len += snprintf(buffer + len, sizeof(buffer) - len, "%s:%d:%d ", hostname, pid, tid); // INFO 级别 len += snprintf(buffer + len, sizeof(buffer) - len, "[%d] NCCL INFO %s\n", cudaDev, fmt);这解释了输出中node-01:1895902:1895913 [0]的含义:主机名、进程 PID、线程 TID 和当前 CUDA 设备号。WARN级别还会额外带上源码位置([cudaDev] 文件:行号 (函数)),并在行首附加一个换行符以便错误更显眼:
if (level == NCCL_LOG_WARN) { buffer[len++] = '\n'; // WARNs come with an extra newline at the beginning. } // ... len += snprintf(buffer + len, sizeof(buffer) - len, "[%d] %s:%d (%s) NCCL WARN %s\n", cudaDev, fileStr, line, funcStr, fmt);另外两个源码层面的细节值得注意:
- 每次 WARN 都会记录最后一条错误。ncclDebugLogV() 在收到 WARN 时会把格式化后的消息存入全局
ncclLastError字符串,供错误码返回后以人类可读形式呈现。 - WARN 后可自动升级为 INFO。src/debug.cc 定义了
NCCL_WARN_ENABLE_DEBUG_INFO参数(默认关闭),启用后一旦记录了一条 WARN 消息,日志级别会自动原子地提升到INFO,方便复现错误前后的上下文。
三、按子系统过滤(NCCL_DEBUG_SUBSYS)
当NCCL_DEBUG=INFO或NCCL_DEBUG=TRACE时,日志量可能非常大。NCCL_DEBUG_SUBSYS允许只保留特定子系统的输出,聚焦关键信息。过滤依据就是上文ncclDebugLogSubSys的位掩码:日志宏携带FLAGS位,与全局ncclDebugMask(由NCCL_DEBUG_SUBSYS解析而来)做与运算,为 0 则丢弃(见 src/include/debug.h 的 INFO 宏)。
基本用法
# 只看网络相关消息 NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=NET ./my_app # 显示多个子系统 NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=INIT,NET,GRAPH ./my_app # 显示除 PROXY 外的所有子系统 NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=^PROXY ./my_app # 排除多个子系统 NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=^PROXY,ALLOC ./my_app以^前缀表示“排除该子系统”(解析逻辑由 src/include/param 下的位集解析器实现)。
可用子系统清单
| 子系统 | 说明 |
|---|---|
| INIT | 初始化与设置 |
| COLL | 集合通信操作(AllReduce、Broadcast 等) |
| P2P | 点对点 send/receive 操作 |
| SHM | 共享内存传输 |
| NET | 网络传输(IB、socket 等) |
| GRAPH | 拓扑检测与图搜索 |
| TUNING | 算法与协议选择 |
| ENV | 环境变量处理 |
| ALLOC | 内存分配 |
| CALL | 函数调用跟踪 |
| PROXY | Proxy 线程操作 |
| NVLS | NVLink SHARP 操作 |
| BOOTSTRAP | 早期初始化与引导(bootstrap) |
| REG | 缓冲区注册 |
| PROFILE | 粗粒度初始化性能剖析 |
| RAS | 可靠性、可用性与可服务性(Reliability, Availability, Serviceability) |
| DESTROY | 通信器销毁、abort、revoke 及插件卸载/关闭操作 |
| ALL | 所有子系统 |
未设置NCCL_DEBUG_SUBSYS时,默认为INIT,BOOTSTRAP,ENV—— 这一点在源码中可以直接印证:src/debug.cc 中ncclParamDebugSubsys的默认值正是NCCL_INIT | NCCL_BOOTSTRAP | NCCL_ENV。此外,源码参数注册表中还存在用户文档表格未单列的ALLOC_HOST(主机内存分配)子系统,见 src/debug.cc。
各子系统的典型输出示例
INIT—— 初始化与插件分配:
node-01:1895902:1895913 [0] NCCL INFO Initialized NET plugin IB node-01:1895902:1895913 [0] NCCL INFO Assigned NET plugin IB to comm node-01:1895902:1895913 [0] NCCL INFO Assigned GIN plugin GIN_XXXX to comm node-01:1895902:1895913 [0] NCCL INFO Assigned RMA plugin RMA_XXXX to comm node-01:1895902:1895913 [0] NCCL INFO Using network IB node-01:1895902:1895913 [0] NCCL INFO DMA-BUF is available on GPU device 0 node-01:1895902:1895913 [0] NCCL INFO [Rank 0] ncclCommInitRankConfig comm 0x... rank 0 nranks 2 cudaDev 0 nvmlDev 0 busId XXXX commId 0x... - Init START node-01:1895902:1895913 [0] NCCL INFO ncclTopoGetCpuAffinity: Affinity for GPU 0 is X,X,X,X,X,X,X,X. (GPU affinity = X,X,X,X,X,X,X,X ; CPU affinity = X-X).GRAPH—— 拓扑检测与通信模式:
node-01:1895811:1895822 [0] NCCL INFO Tree 0 : -1 -> 0 -> 1/-1/-1 node-01:1895811:1895822 [0] NCCL INFO Tree 1 : -1 -> 0 -> 1/-1/-1 node-01:1895811:1895822 [0] NCCL INFO Ring 00 : 1 -> 0 -> 1 node-01:1895811:1895822 [0] NCCL INFO Ring 01 : 1 -> 0 -> 1COLL—— 集合通信操作详情:
node-01:1896279:1896279 [0] NCCL INFO AllReduce: opCount 0 sendbuff 0x... recvbuff 0x... count 2 datatype 7 op 0 root 0 comm 0x... [nranks=2] stream 0x...SHM—— 共享内存传输:
node-01:1896458:1896477 [0] NCCL INFO MMAP allocated shareable host buffer /dev/shm/nccl-XXXXXX size 4096 ptr 0x... node-01:1896458:1896480 [0] NCCL INFO Channel 00 : 0[0] -> 1[1] via SHM/direct/direct node-01:1896458:1896477 [0] NCCL INFO MMAP allocated shareable host buffer /dev/shm/nccl-XXXXXX size 4096 ptr 0x... node-01:1896458:1896480 [0] NCCL INFO Channel 01 : 0[0] -> 1[1] via SHM/direct/directRAS—— 可靠性/可用性/可服务性信息(RAS 子系统的更多介绍见 RAS 文档):
node-01:1896503:1896518 [0] NCCL INFO RAS thread started node-01:1896503:1896518 [0] NCCL INFO RAS handling local addRanks request (old nRasPeers 0) node-01:1896503:1896518 [0] NCCL INFO RAS finished local processing of addRanks request (new nRasPeers 1, nRankPeers 1)DESTROY—— 通信器拆除与插件卸载:
node-01:2056033:2056033 [1] NCCL INFO comm 0x... rank 1 nranks 4 cudaDev 1 busId XXXX - Destroy COMPLETE node-01:2056035:2056035 [3] NCCL INFO comm 0x... rank 3 nranks 4 cudaDev 3 busId XXXX - Destroy COMPLETE node-01:2056032:2056032 [0] NCCL INFO comm 0x... rank 0 nranks 4 cudaDev 0 busId XXXX - Destroy COMPLETE node-01:2056034:2056034 [2] NCCL INFO comm 0x... rank 2 nranks 4 cudaDev 2 busId XXXX - Destroy COMPLETE node-01:2056035:2056035 [3] NCCL INFO ENV/Plugin: Closing env plugin ncclEnvDefault node-01:2056033:2056033 [1] NCCL INFO ENV/Plugin: Closing env plugin ncclEnvDefault node-01:2056034:2056034 [2] NCCL INFO ENV/Plugin: Closing env plugin ncclEnvDefault node-01:2056032:2056032 [0] NCCL INFO ENV/Plugin: Closing env plugin ncclEnvDefaultPROXY—— Proxy 线程操作:
node-01:1896541:1896559 [0] NCCL INFO New proxy recv connection 0 from local rank 0, transport 1 node-01:1896541:1896559 [0] NCCL INFO proxyProgressAsync opId=0x... op.type=1 op.reqBuff=0x... op.respSize=16 done node-01:1896541:1896562 [0] NCCL INFO ncclPollProxyResponse Received new opId=0x... node-01:1896541:1896562 [0] NCCL INFO resp.opId=0x... matches expected opId=0x... node-01:1896541:1896562 [0] NCCL INFO Connected to proxy localRank 0 -> connection 0x... node-01:1896541:1896559 [0] NCCL INFO Received and initiated operation=Init res=0ENV—— 环境变量处理:
node-01:1896826:1896838 [0] NCCL INFO NCCL_SHM_DISABLE set by environment to 1.REG—— 缓冲区注册:
node-01:1896963:1896984 [0] NCCL INFO register comm 0x... buffer 0x... size 8ALLOC—— 内存分配:
node-01:1120933:1120933 [0] NCCL INFO init.cc:2353 Cuda Host Alloc Size 4 pointer 0x... node-01:1120933:1120933 [0] NCCL INFO MemManager: Initialized for device 0 node-01:1120933:1120933 [0] NCCL INFO misc/utils.cc:297 memory stack hunk malloc(65536) node-01:1120933:1120933 [0] NCCL INFO Mem Realloc old size 0, new size 256 pointer 0x... node-01:1120933:1120950 [0] NCCL INFO Mem Realloc old size 0, new size 32 pointer 0x... node-01:1120933:1120933 [0] NCCL INFO channel.cc:43 Cuda CallocAsync Size 608 pointer 0x... memType 2 node-01:1120933:1120933 [0] NCCL INFO channel.cc:46 Cuda CallocAsync Size 24 pointer 0x... memType 2 node-01:1120933:1120933 [0] NCCL INFO channel.cc:58 Cuda CallocAsync Size 4 pointer 0x... memType 2CALL—— 函数调用跟踪(需要NCCL_DEBUG=TRACE):
node-01:1897248:1897248 NCCL CALL ncclGroupStart() node-01:1897248:1897248 NCCL CALL ncclAllReduce(0x...,0x...,2,7,0,0,0x...,0x...) node-01:1897248:1897248 NCCL CALL ncclAllReduce(0x...,0x...,2,7,0,0,0x...,0x...) node-01:1897248:1897248 NCCL CALL ncclGroupEnd()CALL子系统的实现机制是 src/include/debug.h 中的 TRACE_CALL 宏:
#define TRACE_CALL(...) \ do { \ int level = COMPILER_ATOMIC_LOAD(&ncclDebugLevel, std::memory_order_acquire); \ if ((level >= NCCL_LOG_TRACE && (NCCL_CALL & ncclDebugMask)) || (level < 0)) { \ ncclDebugLogInternal(NCCL_LOG_TRACE, NCCL_CALL, nullptr, __func__, __LINE__, __VA_ARGS__); \ } \ } while (0)该宏被嵌入到ncclGroupStart/ncclGroupEnd、ncclSend/ncclRecv等 API 入口处(例如 src/group.cc、src/init.cc、src/enqueue.cc),因此TRACE + CALL组合即可重放一次作业的完整 NCCL API 调用序列。
BOOTSTRAP—— 早期初始化与引导:
node-01:1897326:1897337 [0] NCCL INFO Bootstrap timings total 0.001447 (create 0.000081, send 0.000255, recv 0.000268, ring 0.000068, delay 0.000000)PROFILE—— 粗粒度初始化剖析:
node-01:1897505:1897517 [0] NCCL INFO Bootstrap timings total 0.001204 (create 0.000080, send 0.000303, recv 0.000137, ring 0.000063, delay 0.000000) node-01:1897505:1897517 [0] NCCL INFO Init timings - ncclCommInitRankConfig: rank 0 nranks 2 total 0.37 (kernels 0.28, alloc 0.05, bootstrap 0.00, allgathers 0.00, topo 0.03, graphs 0.00, connections 0.01, rest 0.00)这两条日志分别来自 src/bootstrap.cc 的INFO(NCCL_BOOTSTRAP | NCCL_PROFILE, ...)和 src/init.cc 的初始化耗时统计,可用于定位ncclCommInitRank慢在哪一阶段(kernels、alloc、topo、connections 等)。
NET—— 网络传输操作:
node-01:1897709:1897731 [0] NCCL INFO Connected to proxy localRank 0 -> connection 0x... node-01:1897709:1897731 [0] NCCL INFO Channel 00/0 : 0[0] -> 1[1] [send] via NET/Socket/0 node-01:1897709:1897727 [0] NCCL INFO New proxy send connection 3 from local rank 0, transport 2 node-01:1897709:1897731 [0] NCCL INFO Connected to proxy localRank 0 -> connection 0x... node-01:1897709:1897731 [0] NCCL INFO Channel 01/0 : 0[0] -> 1[1] [send] via NET/Socket/0P2P—— 点对点操作:
node-01:631398:631465 [0] NCCL INFO Allocated shareable buffer 0x... size 10485760 ipcDesc 0x... node-01:631398:631398 [0] NCCL INFO Channel 00/0 : 0[0] -> 1[1] via P2P/CUMEM node-01:631398:631465 [0] NCCL INFO Allocated shareable buffer 0x... size 2097152 ipcDesc 0x... node-01:631398:631398 [0] NCCL INFO Channel 01/0 : 0[0] -> 1[1] via P2P/CUMEMTUNING—— 算法与协议选择:
node-01:631871:631871 [0] NCCL INFO AllReduce: 33554432 Bytes -> Algo RING proto SIMPLE channel{Lo..Hi}={0..31}NVLS—— NVLink SHARP 操作:
node-01:1538174:1538174 [0] NCCL INFO NVLS Creating Multicast group nranks 8 size 2097152 on rank 0 node-01:1538174:1538174 [0] NCCL INFO NVLS Created Multicast group 0x... nranks 8 size 2097152 on rank 0 node-01:1538174:1538174 [0] NCCL INFO NVLS rank 0 (dev 0) alloc done, ucptr 0x... ucgran 2097152 mcptr 0x... mcgran 2097152 ucsize 2097152 mcsize 2097152 (inputsize 24576)四、日志写入文件(NCCL_DEBUG_FILE)
多进程作业中,把所有进程的日志打印到终端会非常混乱。建议用NCCL_DEBUG_FILE为每个进程、每个主机名生成独立日志文件,便于隔离和分析特定 rank 的问题:
NCCL_DEBUG=INFO NCCL_DEBUG_FILE=/tmp/nccl_%h_%p.log ./my_app格式占位符:
%h—— 替换为主机名%p—— 替换为进程 ID(PID)
上例会为每个进程创建形如/tmp/nccl_node-01_12345.log的文件。
注意:文件名模式必须保证进程间唯一(例如包含%p),否则多个进程写同一文件会造成内容交错损坏。
Environment Variables 文档 中还补充了以下适用细节:
- 路径中不能包含
~,需先转换为相对或绝对路径; - 也可以设置为
/dev/stdout或/dev/stderr,将输出定向到预定义 I/O 流,并使其按行缓冲; - 设置
NCCL_DEBUG_FILE会创建文件并覆盖同名旧文件。
源码层面的实现细节
ncclDebugInit() 中的文件处理逻辑:
// 仅当日志级别高于 VERSION 且设置了 NCCL_DEBUG_FILE 时才打开文件 const char* ncclDebugFileEnv = ncclParamDebugFile(); if (tempNcclDebugLevel > NCCL_LOG_VERSION && ncclDebugFileEnv != NULL) { // ... 逐字符解析 %h / %p / %% 占位符 ... FILE* file = fopen(debugFn, "w"); // "w" 模式:截断重写旧文件 if (file != nullptr) { setlinebuf(file); // 禁用块缓冲,按行刷新 ncclDebugFile = file; } }两个关键点:其一,fopen(debugFn, "w")解释了“覆盖同名文件”的行为;其二,setlinebuf(Linux 下;Windows 下使用setvbuf按行缓冲)保证每条日志即时落盘,进程崩溃时不会丢失缓冲中的日志。主机名在替换%h前会经过 getHostNameForLog() 处理,其中的%和/字符会被替换为-,避免破坏路径。
五、时间戳配置
NCCL 日志消息可以带时间戳,用于时序分析。
时间戳格式(NCCL_DEBUG_TIMESTAMP_FORMAT)
使用 strftime 语法自定义格式:
# 默认格式:[YYYY-MM-DD HH:MM:SS] NCCL_DEBUG=INFO NCCL_DEBUG_TIMESTAMP_LEVELS=INFO ./my_app # 包含毫秒 NCCL_DEBUG=INFO NCCL_DEBUG_TIMESTAMP_LEVELS=INFO NCCL_DEBUG_TIMESTAMP_FORMAT="[%F %T.%3f] " ./my_app # 禁用时间戳 NCCL_DEBUG=INFO NCCL_DEBUG_TIMESTAMP_LEVELS=INFO NCCL_DEBUG_TIMESTAMP_FORMAT="" ./my_app默认格式为[%F %T](即[YYYY-MM-DD HH:MM:SS]),定义于 src/debug.cc。除 strftime 转换符外,NCCL 还支持扩展格式%Xf(X 为 1–9 的数字)输出小数秒,例如%3f输出毫秒(零填充),%9f输出纳秒,且整个格式串中只能使用一次。Environment Variables 文档 还提到:格式串中的下划线会被渲染为空格(因为命令行参数中不方便输入空格);时间按本地时区打印,可通过TZ环境变量(如TZ=UTC)调整。
这些行为都能与源码对应:ncclDebugInit() 会扫描格式串寻找%[1-9]f,将其替换为占位空格并在打印时用纳秒值零填充(L304-L315);随后将格式串中的_统一替换为空格。若 strftime 输出空间不足,会自动回退到默认格式[%F %T]。
时间戳级别(NCCL_DEBUG_TIMESTAMP_LEVELS)
控制哪些日志级别带时间戳:
# 为 WARN、INFO、TRACE 都加时间戳 NCCL_DEBUG=INFO NCCL_DEBUG_TIMESTAMP_LEVELS=WARN,INFO,TRACE ./my_app # 除 TRACE 外都加时间戳 NCCL_DEBUG=TRACE NCCL_DEBUG_TIMESTAMP_LEVELS=^TRACE ./my_app默认只有 WARN 级别带时间戳——源码中 ncclParamDebugTimestampLevel 的默认值为(1u << NCCL_LOG_WARN)。可选级别值包括VERSION、WARN、INFO、ABORT、TRACE、ALL,同样支持^排除语法。
需要注意:非 CALL 类的TRACE日志除了可选的 strftime 时间戳外,始终会额外打印自 NCCL 调试子系统初始化以来的微秒数(相对时间)。这一点由 ncclDebugLogV() 中 TRACE 分支 实现:
} else if (level == NCCL_LOG_TRACE) { auto delta = std::chrono::steady_clock::now() - ncclEpoch; double timestamp = std::chrono::duration_cast<...>(delta).count() * 1000; len += snprintf(buffer + len, ..., "[%d] %f %s:%d NCCL TRACE %s\n", cudaDev, timestamp, funcStr, line, fmt); }其中ncclEpoch在 ncclDebugInit() 初始化时取样(std::chrono::steady_clock::now())。因此 TRACE 级别天然具备“可重放”的时间基线,适合与CALL子系统配合分析调用顺序。
六、常见调试场景
1. 诊断初始化挂起
若应用卡在 NCCL 初始化阶段:
NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=INIT,BOOTSTRAP,NET ./my_app关注点:
- Bootstrap 连接问题(Bootstrap timings 是否打印完整)
- 网卡/接口选择问题
- Rank 间同步失败
2. 排查网络问题
针对超时、连接失败等网络相关问题:
NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=NET ./my_app关注点:
- 选择了哪些网卡和网络设备(
Using network IB、Channel xx : ... via NET/...等行) - 连接建立消息
- 网络传输层报出的错误
3. 理解拓扑检测
查看 NCCL 如何探测并使用系统拓扑:
NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=GRAPH ./my_app输出包括:检测到的 GPU 及其互连、NVLink 与 PCIe 拓扑、网络设备 locality 信息,以及最终的 Ring/Tree 通信模式(如Tree 0 : -1 -> 0 -> 1/-1/-1)。
4. 调试性能问题
启用调优信息,观察算法与协议选择:
NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=TUNING ./my_app输出包括:选中的算法(Ring、Tree 等)、协议选择(Simple、LL、LL128)、通道与线程配置。
5. 跟踪 NCCL API 调用
查看所有 NCCL 函数调用:
NCCL_DEBUG=TRACE NCCL_DEBUG_SUBSYS=CALL ./my_app6. 全量调试会话
需要全面排查时,将全部子系统日志按进程分别落盘:
NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=ALL NCCL_DEBUG_FILE=/tmp/nccl_%h_%p.log ./my_app七、实践建议与小结
结合本文档与 env.rst 中的说明,几条实用原则:
- 生产环境保持
NCCL_DEBUG=WARN(或不设置,默认即 WARN),只捕获警告与错误;诊断完成后应移除调试变量,长期开启 INFO/TRACE 会带来明显输出开销。 - 多进程作业务必配合
NCCL_DEBUG_FILE且文件名包含%p,利用每进程独立日志快速定位出问题的 rank;文件按行缓冲,进程异常退出也不会丢日志。 - 优先使用子系统过滤缩小日志范围:初始化问题看
INIT,BOOTSTRAP,NET,传输层问题看NET或SHM,性能问题看TUNING,全量排查才用ALL。 - 需要时序分析时用时间戳:
NCCL_DEBUG_TIMESTAMP_LEVELS=INFO加NCCL_DEBUG_TIMESTAMP_FORMAT="[%F %T.%3f] "可获得毫秒级绝对时间戳;TRACE 级别自带相对微秒时钟,适合与CALL跟踪配合重放调用序列。 - 善用 WARN 的源码定位信息:WARN 行会携带
文件:行号 (函数)(如plugin/net.cc:334),可直接在当前仓库中检索对应代码(例如 plugins/net/example/plugin.c 所在插件链路),快速定位报错源头。
NCCL 日志系统的完整参考可见仓库中的 Logging 文档 与 Environment Variables 文档;核心实现则集中在 src/debug.cc、src/include/debug.h 与 src/include/nccl_common.h 三个文件中,本文所有级别、子系统、默认值与格式行为均以上述源码为准。
- 高性能计算
- 通信
- 深度学习
- 分布式训练
【免费下载链接】nccl
Optimized primitives for collective multi-GPU communication
相关推荐
Flecs日志系统:自定义日志输出与过滤
Flecs日志系统:自定义日志输出与过滤 你是否在调试Flecs应用时被大量日志淹没?是否需要将关键错误信息定向到特定监控系统?本文将详细介绍如何通过Flecs
游戏开发Zephyr RTOS调试日志时间戳:系统时间同步
Zephyr RTOS调试日志时间戳:系统时间同步 你是否在调试嵌入式设备时遇到过日志时间混乱、事件顺序难以追踪的问题?Zephyr RTOS(Real Tim
操作系统嵌入式RTOS物联网Kedro 日志系统实战指南:从默认配置到 Rich 格式化、文件日志与自定义过滤
Kedro 日志系统实战指南:从默认配置到 Rich 格式化、文件日志与自定义过滤 导读 日志是数据管道可观测性的基石。Kedro 直接基于 Python 标准
数据工程工作流自动化
创作声明:本文部分内容由AI辅助生成(AIGC),仅供参考