技术开发 频道

CUDA Texture Part.2 Linear Memory

  【IT168 文档】在前一篇已經大概介紹過 texture 的基本概念了,在這一篇,就來大概看一下,一般使用 linear memory 的 texture 的方法吧。

  在宣告的部分,前面已經提過了,大致上就是:

texture<int, 1, cudaReadModeElementType> texRef;

   的形式。這樣,就可以宣告出一個一維的整數 texture 了;而由於他的讀取模式是給定為 cudaReadModeElementType,所以之後由這個 texture 取出的值,也都會是整數(如果是 cudaReadModeNormalizedFloat,就會是浮點數)。

  而宣告出 texture reference 的物件後,接下來就是要和現有的變數做連結了~

  Bind Texture

  使用 linear memory 時,是要將在 global memory 中用 cudaMalloc() 定義出來的記憶體空間,透過 cudaBindTexture() 這個函式,來將 texture reference 連結到變數。其函式的型態為:

template<class T, int Dim, enum cudaReadModeElementType readMode>
cudaError_t cudaBindTexture( size_t* offset,
                         const struct texture<T, dim, readMode>& texRef,
                         const void* devPtr,
                         size_t size = UINT_MAX);

   其中,offset 算是比較進階的設定,在這邊先略過,不過一般是都給 0。而 texRef 就是要 bind 的 texture,devPtr 則是要 bind 到 texRef 的資料;size 就是 devPtr 的記憶體空間大小,不過一般應該是可以省略不給。

  而除了 cudaBindTexture() 外,當然也有相反的 cudaUnbindTexture()。他的動作就是解除 texture reference 和變數的關係。

template<class T, int Dim, enum cudaReadModeElementType readMode>
cudaError_t cudaunBindTexture(
                const struct texture<T, dim, readMode>& texRef );

   在 kernel 中使用

  在 kernel function 中,要存取使用 linear memory 的 texture reference,要透過 tex1Dfetch() 這個函式;其型態為:

template<class Type>Type tex1Dfetch( texture<Type, 1, ReadMode> texRef, int x);

   也就是只要給他要讀取的 texture reference,以及要讀取的位置 x,就可以取得資料的值了~

0
相关文章