分析一个CUDA矩阵加法代码,使用nvprof:代码API配置文件,内核没有

Posted

技术标签:

【中文标题】分析一个CUDA矩阵加法代码,使用nvprof:代码API配置文件,内核没有【英文标题】:Profiling a CUDA matrix addition code, using nvprof: the code API profiles, the kernel does not 【发布时间】:2021-12-07 15:54:09 【问题描述】:

我正在使用带有 nvidia Geforce Gpu 的远程工作站,并且在编译和执行后,当我尝试分析时,它会显示在屏幕上

这是我运行 nvidia-smi 时的输出

#include <stdio.h>
#include <cuda.h>
#include <math.h>

__global__ void matrixInit(double *matrix, int width, int height, double value)
    for(int i = (threadIdx.x + blockIdx.x * blockDim.x); i<width; i+=(blockDim.x * gridDim.x))
        for(int j = (threadIdx.y + blockIdx.y * blockDim.y); j<height; j+=(blockDim.y * gridDim.y))
            matrix[j * width +i] = value;
        
    


__global__ void matrixAdd(double *d_A, double *d_B, double *d_C, int width, int height)
    int ix = threadIdx.x + blockIdx.x * blockDim.x;
    int iy = threadIdx.y + blockIdx.y * blockDim.y;

    int stride_x = blockDim.x * gridDim.x;
    int stride_y = blockDim.y * gridDim.y;

    for(int j=iy; j<height; j+=stride_y)
        for(int i=ix; i<width; i+=stride_x)
            int index = j * width +i;
           d_C[index] = d_A[index-1] + d_B[index];
        
    


int main()
    int Nx = 1<<12;
    int Ny = 1<<15;


    size_t size = Nx*Ny*sizeof(double);

 // host memory pointers
    double *A, *B, *C;

 // device memory pointers
    double *d_A, *d_B, *d_C;

    // allocate host memory
    A = (double*)malloc(size);
    B = (double*)malloc(size);
    C = (double*)malloc(size);

    // kernel call
    int thread = 32;
    int block_x = ceil(Nx + thread -1)/thread;
    int block_y = ceil(Ny + thread -1)/thread;

    dim3 THREADS(thread,thread);
    dim3 BLOCKS(block_y,block_x);

    // initialize variables
    matrixInit<<<BLOCKS,THREADS>>>(A, Nx, Ny, 1.0);
    matrixInit<<<BLOCKS,THREADS>>>(B, Nx, Ny, 2.0);
    matrixInit<<<BLOCKS,THREADS>>>(C, Nx, Ny, 0.0);

    //allocated device memory

    cudaMalloc(&d_A, size);
    cudaMalloc(&d_B, size);
    cudaMalloc(&d_C, size);


//copy to device
    cudaMemcpy(d_A, A, size, cudaMemcpyHostToDevice);
    cudaMemcpy(d_B, B, size, cudaMemcpyHostToDevice);


// Add matrix at GPU
    matrixAdd<<<BLOCKS,THREADS>>>(A, B, C, Nx, Ny);

//copy back to host
    cudaMemcpy(C, d_C, size, cudaMemcpyDeviceToHost);

    cudaFree(A);
    cudaFree(B);
    cudaFree(C);

    return 0;


这是我的代码。总之,结果显示以下两条警告消息:

==525867== Warning: 4 records have invalid timestamps due to insufficient device buffer space. You can configure the buffer space using the option --device-buffer-size.                
==525867== Warning: 1 records have invalid timestamps due to insufficient semaphore pool size. You can configure the pool size using the option --profiling-semaphore-pool-size. 
==525867== Profiling result: No kernels were profiled.

【问题讨论】:

在 your previous question 上,我要求提供来自 nvidia-smi 的输出(我的意思不仅仅是该输出的前 15 个字符),我还建议您添加 proper CUDA error checking。 哦,是的。谢谢。固定的。我添加了 Nvidia-smi 命令屏幕截图的输出。我也改进了代码。 您的代码中没有正确的 CUDA 错误检查。而且我无法从 nvidia-smi 输出中确定您拥有哪种 GPU。请改为运行以下命令:nvidia-smi -q |grep -i name 并提供该命令的结果。 我很抱歉我是 cuda 编程的新手。你的命令返回:NVIDIA GeForce GTX 980 【参考方案1】:
matrixInit<<<BLOCKS,THREADS>>>(A, Nx, Ny, 1.0);
matrixInit<<<BLOCKS,THREADS>>>(B, Nx, Ny, 2.0);
matrixInit<<<BLOCKS,THREADS>>>(C, Nx, Ny, 0.0);

你在这里写入主机内存,这是不允许的。

相反,您可以在分配后直接在设备阵列d_Ad_Bd_C 上执行matrixInit()

这里的另一个错误:

cudaFree(A);
cudaFree(B);
cudaFree(C);

应该是d_Ad_Bd_C。对ABC 使用常规的free()

你的内核也没有做你想做的事。每个矩阵条目都使用一个线程启动它们,这意味着内核中不应有 for() 循环。

【讨论】:

以上是关于分析一个CUDA矩阵加法代码,使用nvprof:代码API配置文件,内核没有的主要内容,如果未能解决你的问题,请参考以下文章

如何仅通过 nvprof 分析 CUDA 应用程序

使用 CUDA Profiler nvprof 进行内存访问

在 CUDA 分析期间 <overflow> 是啥意思?

如何让 malloc 出现在 nvprof 的统计分析器中?

Nvidia 视觉分析器错误:无法创建分析文件

[CUDA]CUDA编程实战三——矩阵加法的实现