CUDA纹理记忆空间

pQB*_*pQB 8 memory cuda hierarchy texture-mapping

当我将数组绑定到CUDA中的纹理时,

  1. 那个数组复制到纹理空间?要么,
  2. 是作为纹理的数组引用?

如果答案是1.,那么我可以从纹理存储空间绑定纹理和安全获取数据,同时将结果写入数组,该数组在全局存储器中分配.

如果答案是2.那么纹理存储器是一个全局存储空间,数据被缓存并空间获取?

我想知道这个话题,因为我已经看到了一些与这个话题相关的问题,我现在还没有明确答案.

提前致谢.

tal*_*ies 16

答案是第二种选择,但从那里开始变得更加复杂.没有"纹理存储器"这样的东西,只有通过专用硬件访问的全局存储器,其中包括GPU读取缓存(每个MP 6-8kb,具体取决于卡,参见Cuda编程指南附录F中的表F-2)以及许多硬件加速过滤/插值动作.纹理硬件可以通过两种方式在CUDA中使用:

  1. 将线性内存绑定到纹理,并使用1D fetch API在内核中读取它.在这种情况下,纹理硬件实际上只是作为直读缓存,而(IIRC)没有可用的过滤操作.
  2. 创建一个CUDA数组,将线性内存的内容复制到该数组,并将其绑定到纹理.生成的CUDA数组包含线性源的空间有序版本,存储在某种(未记录的)空间填充曲线的全局存储器中.纹理硬件提供对该阵列的缓存访问,包括使用硬件加速过滤的同时内存读取.

您可能会发现David Kanter编写的GT200体系结构概述值得一读,以便更好地了解实际体系结构如何实现API公开的内存层次结构.

  • 在CUDA中,纹理和数组是不透明的只读对象.不允许写入(甚至不保证内部纹理布局在不同硬件之间是相同的).如果您有Fermi卡并且正在使用CUDA 3.2或更高版本,则可以使用曲面.虽然没有过滤支持,[表面API](http://developer.download.nvidia.com/compute/cuda/3_1/toolkit/docs/online/group__CUDART__SURFACE.html)提供了对"纹理"的读写操作.目前,AFIAK. (3认同)
  • 在内核调用中,纹理缓存不会与底层全局内存存储保持一致,请参阅编程指南的第3.2.10.4节.使用过滤的2D纹理不必绑定到不透明的cudaArray,它们也可以绑定到使用cudaMallocPitch()分配的音高线性内存. (3认同)