gpt4 book ai didi

optimization - 为什么 NVRTC 不优化我的整数除法和模运算?

转载 作者:行者123 更新时间:2023-12-03 16:56:49 24 4
gpt4 key购买 nike

我在 NVRTC 编译了一个内核:

__global__ void kernel_A(/* args */) {
unsigned short idx = threadIdx.x;
unsigned char warp_id = idx / 32;
unsigned char lane_id = idx % 32;
/* ... */
}

我知道整数除法和模运算在 CUDA GPU 上的成本非常高。但是我认为这种 2 的幂除法应该优化为位运算,直到我发现它不是:

__global__ void kernel_B(/* args */) {
unsigned short idx = threadIdx.x;
unsigned char warp_id = idx >> 5;
unsigned char lane_id = idx & 31;
/* ... */
}

kernel_B 似乎运行得更快。当忽略内核中的所有其他代码时,以 1024 个大小为 1024 的 block 启动时,nvprof 显示 kernel_A 平均运行 15.2us,而 kernel_B 平均运行 7.4us。我推测 NVRTC 没有优化整数除法和模数。

结果是在 GeForce 750 Ti、CUDA 8.0 上获得的,取 100 次调用的平均值。提供给 nvrtcCompileProgram() 的编译器选项是 -arch compute_50

这是预期的吗?

最佳答案

在代码库中进行了彻底的错误扫描。原来我的应用程序是在 DEBUG 模式下构建的。这会导致额外的标志 -G-lineinfo 传递给 nvrtcCompileProgram()

来自 nvcc 手册页:

--device-debug (-G)

Generate debug information for device code. Turns off all optimizations. Don't use for profiling; use -lineinfo instead.

关于optimization - 为什么 NVRTC 不优化我的整数除法和模运算?,我们在Stack Overflow上找到一个类似的问题: https://stackoverflow.com/questions/44300294/

24 4 0
Copyright 2021 - 2024 cfsdn All Rights Reserved 蜀ICP备2022000587号
广告合作:1813099741@qq.com 6ren.com