【问题标题】:Count the number of cycles in a CUDA kernel计算 CUDA 内核中的周期数
【发布时间】:2013-05-19 01:09:10
【问题描述】:

如何计算如下函数执行的周期数。我应该直接计算 sum 和 muls 和 divs 的数量吗?在哪里可以查看 CUDA 中的加法需要多少个周期?

__global__
void mandelbrotSet_per_element(Grayscale *image){
    float minR = -2.0f, maxR = 1.0f;
    float minI = -1.2f, maxI = minI + (maxR-minR) * c_rows / c_cols;
    float realFactor = (maxR - minR) / (c_cols-1);
    float imagFactor = (maxI - minI) / (c_rows-1);

    bool isInSet;
    float c_real, c_imag, z_real, z_imag;

    int y = blockDim.y * blockIdx.y + threadIdx.y;
    int x = blockDim.x * blockIdx.x + threadIdx.x;

    while (y < c_rows){
        while (x < c_cols) {
            c_real = minR + x * realFactor;
            c_imag = maxI - y * imagFactor;
            z_real = c_real;    z_imag = c_imag;
            isInSet = true;

            for (int k = 0; k < c_iterations; k++){
                float z_real2 = z_real * z_real;
                float z_imag2 = z_imag * z_imag;
                if (z_real2 + z_imag2 > 4){
                    isInSet = false;
                    break;
                }
                z_imag = 2 * z_real * z_imag + c_imag;
                z_real = z_real2 - z_imag2 + c_real;
            }
            if (isInSet)    image[y*c_cols+x] = 255;
            else            image[y*c_cols+x] = 0;

            x += blockDim.x * gridDim.x;
        }
        x = blockDim.x * blockIdx.x + threadIdx.x;
        y += blockDim.y * gridDim.y;
    }
}

【问题讨论】:

    标签: cuda gpgpu


    【解决方案1】:

    编程指南here中描述了指令吞吐量

    您还可以尝试使用 here 描述的本机 clock() 函数测量一系列指令

    编译器往往会掩盖源代码级别的实际操作计数(增加或可能减少明显的算术强度),因此如果您想准确识别机器正在做什么,您可能需要检查 ptx (nvcc -ptx . ..) 或者可能是机器汇编级代码,称为 SASS,您可以使用 cuobjdump 实用程序从可执行文件中提取。

    【讨论】:

    • 非常感谢。因此,例如,如果内核进行 8 次加法,则所需的周期数是 8/32 ?根据指令吞吐量?
    • 如果内核在单个线程中进行 8 SP FP 添加,即按顺序,在 CC 2.0 设备上运行,并忽略其他因素,例如 ILP、数据停顿、寄存器争用、愚蠢的编译器技巧等,那么它应该需要 8 个时钟,或者更准确地说,SM 可以每时钟退出 1 个添加(这就是 throughput 的含义)。如果一个 warp 中的所有线程都在执行相同的 8 个加法,则不再需要(具有相同的警告)。如果 warp 中只有 8 个线程在执行相同的 8 个加法,则所花费的时间不会少。
    猜你喜欢
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 1970-01-01
    • 2019-06-14
    • 2016-02-21
    • 1970-01-01
    相关资源
    最近更新 更多