6

我正在学习 cuda 纹理记忆。现在，我得到了一个 opencv Iplimage，并得到了它的图像数据。然后我将纹理绑定到这个 uchar 数组，如下所示：

Iplimage *image = cvCreateImage(cvSize(width, height), IPL_DEPTH_8U, 3);
unsigned char* imageDataArray = (unsigned char*)image->imagedata;

texture<unsigned char,2,cudaReadModeElementType> tex;
cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc(8, 8, 8, 0, 
                                                          cudaChannelFormatKindUnsigned); 
cudaArray *cuArray = NULL;
CudaSafeCall(cudaMallocArray(&cuArray,&channelDesc,width,height));

cudaMemcpy2DToArray(cuArray,0,0,imageDataArray,image->widthstep,
    width * sizeof(unsigned char), height, cudaMemcpyHostToDevice);
cudaBindTextureToArray(texC1_cf,cuArray_currentFrame, channelDesc);

现在我启动内核，我想访问该图像的每个像素，每个通道。这就是我感到困惑的地方。

我使用此代码获取像素坐标（X，Y）：

int X = (blockIdx.x*blockDim.x+threadIdx.x);
int Y = (blockIdx.y*blockDim.y+threadIdx.y);

我怎样才能访问这个（X，Y）的每个通道？下面的代码返回什么？

tex2D(tex, X, Y);

除此之外，你能告诉我纹理内存是如何使用纹理来访问数组的，以及这个变换是什么样子的吗？

在此处输入图像描述

4

1 回答 1

5

要将 3 通道 OpenCV 图像绑定到 cudaArray 纹理，您必须创建一个宽度等于 的 cudaArray image->width * image->nChannels，因为通道是由 OpenCV 交错存储的。

cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<unsigned char>();

cudaArray *cuArray = NULL;
CudaSafeCall(cudaMallocArray(&cuArray,&channelDesc,width * image->nChannels,height));

cudaMemcpy2DToArray(cuArray,0,0,imageDataArray,image->widthstep, width * image->nChannels * sizeof(unsigned char), height, cudaMemcpyHostToDevice);

cudaBindTextureToArray(texC1_cf,cuArray_currentFrame, channelDesc);

现在，要在内核中分别访问每个通道，您只需将 x 索引乘以通道数并添加所需通道的偏移量，如下所示：

unsigned char blue = tex2D(tex, (3 * X) , Y);
unsigned char green = tex2D(tex, (3 * X) + 1, Y);
unsigned char red = tex2D(tex, (3 * X) + 2, Y);

第一个是蓝色的，因为 OpenCV 存储具有通道序列 BGR 的图像。

至于您尝试使用访问时遇到的texture<uchar3,..>错误tex2D；CUDA 仅支持创建 1,2 和 4 元素矢量类型的 2D 纹理。不幸的是，ONLY 3 不受支持，这对于绑定 RGB 图像非常有用，并且是一个非常理想的功能。

于 2013-04-25T14:31:44.867 回答