☰
NCCL 日志系统完全指南:NCCL_DEBUG、子系统过滤、日志文件与时间戳配置
2026/10/13 21:55:52 网站建设 项目流程
  • 高性能计算
  • 通信
  • 深度学习
  • 分布式训练

【免费下载链接】nccl

Optimized primitives for collective multi-GPU communication

项目地址:https://gitcode.com/gh_mirrors/nc/nccl
点击查看免费下载

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.0

NCCL_DEBUG=WARN—— 打印警告与错误:

[2026-05-05 06:16:47] node-01:189884:189884 [3] plugin/net.cc:334 NCCL WARN Failed to initialize any NET plugin

NCCL_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 0

NCCL_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);

另外两个源码层面的细节值得注意:

  1. 每次 WARN 都会记录最后一条错误。ncclDebugLogV() 在收到 WARN 时会把格式化后的消息存入全局ncclLastError字符串,供错误码返回后以人类可读形式呈现。
  2. 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函数调用跟踪
PROXYProxy 线程操作
NVLSNVLink 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 -> 1

COLL—— 集合通信操作详情:

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/direct

RAS—— 可靠性/可用性/可服务性信息(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 ncclEnvDefault

PROXY—— 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=0

ENV—— 环境变量处理:

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 8

ALLOC—— 内存分配:

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 2

CALL—— 函数调用跟踪(需要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/0

P2P—— 点对点操作:

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/CUMEM

TUNING—— 算法与协议选择:

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_app

6. 全量调试会话

需要全面排查时,将全部子系统日志按进程分别落盘:

NCCL_DEBUG=INFO NCCL_DEBUG_SUBSYS=ALL NCCL_DEBUG_FILE=/tmp/nccl_%h_%p.log ./my_app

七、实践建议与小结

结合本文档与 env.rst 中的说明,几条实用原则:

  1. 生产环境保持NCCL_DEBUG=WARN(或不设置,默认即 WARN),只捕获警告与错误;诊断完成后应移除调试变量,长期开启 INFO/TRACE 会带来明显输出开销。
  2. 多进程作业务必配合NCCL_DEBUG_FILE且文件名包含%p,利用每进程独立日志快速定位出问题的 rank;文件按行缓冲,进程异常退出也不会丢日志。
  3. 优先使用子系统过滤缩小日志范围:初始化问题看INIT,BOOTSTRAP,NET,传输层问题看NET或SHM,性能问题看TUNING,全量排查才用ALL。
  4. 需要时序分析时用时间戳:NCCL_DEBUG_TIMESTAMP_LEVELS=INFO加NCCL_DEBUG_TIMESTAMP_FORMAT="[%F %T.%3f] "可获得毫秒级绝对时间戳;TRACE 级别自带相对微秒时钟,适合与CALL跟踪配合重放调用序列。
  5. 善用 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

项目地址:https://gitcode.com/gh_mirrors/nc/nccl
点击查看免费下载

相关推荐

创作声明:本文部分内容由AI辅助生成(AIGC),仅供参考

需要专业的网站建设服务?

联系我们获取免费的网站建设咨询和方案报价,让我们帮助您实现业务目标

立即咨询