-4

我正在尝试在结构中使用and 在reqRegs运行时动态优化内核的块大小。sharedSizeBytescudaFuncAttributes

我当前的实现从标准输出文本中浏览nvcc --ptxas-options=-v以发现内核的寄存器和共享内存使用情况。这种方法有点hacky，并且取决于输出文本的确切格式--ptxas-options=-v，可能会在没有警告的情况下更改。

我的问题是我看到--ptxas-options=-v输出中报告的“smem”共享内存值sharedSizeBytes与cudaFuncAttributes结构中的差异，这让我担心我一直使用的共享内存估计是错误的，或者sharedSizeBytes变量不可靠，这意味着我不能将它用于运行时块大小优化。这是nvcc --ptxas-options=-v一个这样的内核的输出......

ptxas info    : Used 14 registers, 2088 bytes smem, 48 bytes cmem[1]

...与cudaFuncAttributes.sharedSizeBytes运行时的值 = 296 相比，对于完全相同的内核。有人知道这里会发生什么吗？

这是另一个使用不同内核的示例：

ptxas info    : Used 18 registers, 2132 bytes smem, 48 bytes cmem[1]

其中cudaFuncAttributes.sharedSizeBytes= 340 在运行时。

谢谢。

4

1 回答 1

0

谢谢罗伯特和马可的回复。他们帮助我排除了一些情况。

事实证明，报告的共享内存使用量不匹配是由于第一次测试编译后使用的共享内存量（由报告）--ptxas-options=-v与修改后的块大小的最终程序使用的共享内存量不同（由 报告cudaFuncAttributes.sharedSizeBytes） . （为清楚起见进行编辑）

共享内存的差异是由于共享内存数组分配依赖于块大小造成的；例如：

__shared__ float myArray[BLOCK_SIZE];

上面的语句在块大小为 256 的程序中使用不同数量的共享内存，而不是使用优化块大小 192 编译的相同源代码。现在看起来很明显，但在优化的 CUDA 代码中需要注意一些事情一代。

于 2014-07-14T13:43:34.113 回答