问题标签 [bank-conflict]

可以使用 nvprof 访问/读取 CUDA exec 的银行冲突计数器：

但是，它不适用于使用 OpenCL 而不是 CUDA 代码的代码。

• 有没有办法nvprof从 OpenCL 环境中提取这些计数器，可能直接从 ptx 中提取？
• 或者，有什么方法可以将使用clGetProgramInfowith的 nvidia OpenCL 编译器生成的 PTX 程序集转换CL_PROGRAM_BINARIES为 CUDA 内核并使用它运行它cuModuleLoadDataEx，从而能够使用nvprof？
• 是否有任何模拟 CPU 后端允许设置诸如银行大小等参数？

附加选项：

• 使用 opencl 到 cuda 代码的转换器，包括 CUDA 中缺少的功能，如 vloadn/vstoren、float16 和其他各种访问器。#define仅适用于简单内核。有没有提供它的工具？

我想确保我正确理解共享内存中的银行冲突。我有 32 部分数据。这些部分由 128 个整数组成。

|0, 1, 2, ..., 125, 126, 127| ... |3968, 3969, 3970, ..., 4093, 4094, 4095|

经线中的每个线程只能访问它自己的部分。

• 线程 0 访问第 0 部分中的位置 0(0)

• 第 1 部分中的线程 1 访问位置 0(128)

• 线程 31 访问第 31 部分中的位置 0(3968)

这是否意味着我这里有 32 个冲突？如果是，那么如果我将部分拉伸到 129 个元素，那么每个线程将访问唯一的银行。我对吗？

我正在编写一个 Cuda 程序。共享内存的大小为__shared__ int shared_memory[1024]每块 32x32 个线程。第一轮，每个线程访问shared_memory[0]中的值并进行计算，第二轮，每个线程访问shared_memory[31]中的值，...，第32轮，每个线程将访问 shared_memory[1023] 中的值。

我想知道是否有一种有效的方法可以避免线程在每一轮中访问相同的共享内存位置？我知道内存填充是一种避免银行冲突的方法，但是程序需要访问共享内存中的相同位置。

我对银行冲突有点困惑，避免使用内存填充和合并内存访问。到目前为止我读过的内容：从全局内存中合并内存访问是最佳的。如果无法实现，则可以使用共享内存对当前块所需的数据进行重新排序，从而使合并访问成为可能。但是，当使用共享内存时，必须注意银行冲突。避免存储库冲突的一种策略是将存储在共享内存中的数组填充 1。考虑此博客文章中的示例，其中 16x16 矩阵的每一行都填充 1，使其成为共享内存中的 16x17 矩阵。

现在我明白了使用内存填充可能会避免银行冲突，但这是否也意味着内存不再对齐？例如，如果我将全局内存移动 1 从而使其未对齐，则一个扭曲将需要访问两个内存通道而不是一个，因为最后一个数字与所有其他数字不在同一通道中。因此，据我了解，合并的内存访问和内存填充是相互矛盾的概念，不是吗？一些澄清非常感谢！

如果同一块中的所有线程访问相同的地址，即用于旧功能的数组 [0]，存在存储库冲突，但对于最新功能（即 GPU V100 为 7.0 或 A100 为 8.0）仍然存在此冲突？