2017-01-02 60 views
1

我想用CUDA处理图像。每个像素的新值是基于一行中的两个相邻像素计算的。对于像素值使用__shared__内存是否有意义,因为每个值只能使用两次?不是瓷砖也是错误的做法,因为它不适合问题结构?我的方法是在每个像素上运行一个线程,并为每个线程每次加载相邻的像素值。将共享内存用于相邻的数组元素?

+0

谁投票结束这个问题?你可以在下面看到,这个问题不是太宽泛,可以简短地回答它。虽然我同意现在很多CUDA问题的质量都不高,但这并不是,我认为立即关闭问题的习惯已经变得非常常见(我会很高兴地承认这包括我自己)。 – tera

回答

4

目前支持的所有CUDA架构都有缓存。 从计算能力3.5开始,这些对于只读数据特别有效(因为读写数据只能在L2中缓存,L1缓存仅限于只读数据)。如果您将输入数据的指针标记为const __restrict__,则编译器很可能会加载它via the L1 texture cache。您也可以通过明确使用__ldg() builtin来强制执行此操作。

虽然可以通过共享内存显式管理从邻近像素重新使用数据,但您可能会发现这比只依赖缓存没有什么好处。

当然,无论您是否使用共享内存,您希望在x方向上最大化块大小,并使用1的blockSize.y作为最佳访问局部性。

1

结合使用共享内存和利用合并内存访问。你所需要做的就是确保图像以行方式存储。每个块都会处理一个线性数组块。由于数据重用(除了第一个和最后一个像素都会参与处理三次),如果在内核开始时您将复制将要处理的所有像素的值到共享内存,将是有益的。