1，编译

下载源代码

git clone --recursive https://github.com/NVIDIA/nccl-tests.git

 

编译源代码

cd nccl-tests/
make -j

2，运行
 

cd ./build/
./all_reduce_perf --help
./all_reduce_perf -b 8 -e 256M -f 2 -g 4

效果图：

3，注意

在这些testcase中，需要注意这两个函数：

这两函数配合出现，起到了一个类似大括号的作用{ ... }

两者之间调用的 nccl API 函数组成要给 group，表示这是一组关系紧密的函数操作。

ncclGroupStart() 调用后，接下来的nccl API 调用都会是非阻塞于GPU 操作的方式立即返回；

ncclGroupEnd() 返回时，仅仅表示已经将nccl API所引发的 cuda gpu 等相关的操作已经提交排队进入了对应的 cuda stream里了，但并不能保证gpu内部已经执行完毕，用cuda 编程的角度看，也就是相关的 cuda kernel 已经全部 被 launch 起来了，但不一定开始执行了，更不一定执行完毕了；

所以，通常再ncclGroupEnd()后面还会调用一个

cudaStreamSynchronize(...);

这是一个阻塞于gpu操作的cuda API，只有等到这个cuda stream中的之前提交的 gpu 操作都做完后才会返回。

参考：

Group Calls — NCCL 2.19.3 documentation