时间戳

首页 > 解决方案 > NCCL 项目的 NVProf

cuda - NCCL 项目的 NVProf

问题描述

当我想将 NVProf 用于带有 --metrics all 的 NCCL 问题时，分析结果总是像我一样返回

==2781== NVPROF is profiling process 2781, command: ./nccl_example 2 16
==2781== Profiling application: ./nccl_example 2 16
==2781== Profiling result:

没有分析事件/指标。我的简单 nccl 程序

#include <stdio.h>
#include "cuda_runtime.h"
#include "nccl.h"

int main(int argc, char* argv[])
{
  ncclComm_t comms[4];


  // managing 4 devices
  int nDev = 3;
  int size = 32*1024*1024;
  int devs[4] = {0, 1, 2};

  //allocating and initializing device buffers
  float** sendbuff = (float**)malloc(nDev * sizeof(float*));
  float** recvbuff = (float**)malloc(nDev * sizeof(float*));
  cudaStream_t* s = (cudaStream_t*)malloc(sizeof(cudaStream_t)*nDev);


  for (int i = 0; i < nDev; ++i) {
    CUDACHECK(cudaSetDevice(i));
    CUDACHECK(cudaMalloc(sendbuff + i, size * sizeof(float)));
    CUDACHECK(cudaMalloc(recvbuff + i, size * sizeof(float)));
    CUDACHECK(cudaMemset(sendbuff[i], 1, size * sizeof(float)));
    CUDACHECK(cudaMemset(recvbuff[i], 0, size * sizeof(float)));
    CUDACHECK(cudaStreamCreate(s+i));
  }


  //initializing NCCL
  NCCLCHECK(ncclCommInitAll(comms, nDev, devs));

   //calling NCCL communication API. Group API is required when using
   //multiple devices per thread
  NCCLCHECK(ncclGroupStart());
  for (int i = 0; i < nDev; ++i)
    NCCLCHECK(ncclAllReduce((const void*)sendbuff[i], (void*)recvbuff[i], 
                                size, ncclFloat, ncclSum, comms[i], s[i]));
  NCCLCHECK(ncclGroupEnd());


  //synchronizing on CUDA streams to wait for completion of NCCL operation
  for (int i = 0; i < nDev; ++i) {
    CUDACHECK(cudaSetDevice(i));
    CUDACHECK(cudaStreamSynchronize(s[i]));
  }


  //free device buffers
  for (int i = 0; i < nDev; ++i) {
    CUDACHECK(cudaSetDevice(i));
    CUDACHECK(cudaFree(sendbuff[i]));
    CUDACHECK(cudaFree(recvbuff[i]));
  }


  //finalizing NCCL
  for(int i = 0; i < nDev; ++i)
      ncclCommDestroy(comms[i]);

  printf("Success \n");
  return 0;
}

因为我需要了解 NCCL API 的详细指标，这样我才能更深入地了解它的性能。

标签： cudanvidia

解决方案

这种行为是意料之中的。

默认情况下收集的事件、指标与 CUDA 设备代码活动有关。要查看可能有启发性的内容，请尝试使用--print-gpu-traceswitch （和 remove --metrics all）进行分析。

记录在案的 “指标”不适用于 NCCL 正在执行的操作（数据复制）。它们适用于 CUDA 内核（即 CUDA 设备代码活动）。

nvprof似乎确实有可以为 NVLink 活动收集的指标。要查看这些，请在适用的系统（例如具有 NVLink）上运行以下命令：

nvprof --query-metrics

或者

nvprof --query-metrics |grep -i nvlink

推荐阅读