2

次のように、C側に1チャンネルのフロート画像があります。

int width, height;
float* img;

この画像を CUDA テクスチャに渡したいです。NVIDIA CUDA C プログラミング ガイド(ページ 42-43) を読み、チュートリアルを使用して、次のようなコードを書きました。

main.cpp:

int main()
{
     int width, height;
     float* h_Input;
     ReadImage(&h_Input, &width, &height); // My function which reads the image.
     WriteImage(h_Input, width, height); // works perfectly...

     float* h_Output = (float*) malloc(sizeof(float) * width * height);

     CalculateWithCuda(h_Input, h_Output, width,height);
     WriteImage(h_Output, width, height); // writes an empty-gray colored image.... *WHY???* 
}

kernel.cu:

texture<float, cudaTextureType2D, cudaReadModeElementType> texRef; // 2D float texture

__global__ void Kernel(float* output, int width, int height)
{
    int i = blockIdx.y * blockDim.y + threadIdx.y; // row number 
    int j = blockIdx.x * blockDim.x + threadIdx.x; // col number

    if(i < height && j < width)
    {
           float temp = tex2D(texRef, i + 0.5f, j + 0.5f);
           output[i * width + j] = temp ;
    }
} 

void CalculateWithCuda(const float* h_input, float* h_output, int width, int height)
{
    float* d_output;

    // Allocate CUDA array in device memory
    cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc(32, 0, 0, 0,cudaChannelFormatKindFloat);
    cudaArray* cuArray;
    cudaMallocArray(&cuArray, &channelDesc, width, height);
    // Copy to device memory some data located at address h_data in host memory 
    cudaMemcpyToArray(cuArray, 0, 0, h_input, width * height * sizeof(float) , cudaMemcpyHostToDevice);
    // Set texture parameters
    texRef.addressMode[0] = cudaAddressModeWrap;
    texRef.addressMode[1] = cudaAddressModeWrap;
    texRef.filterMode     = cudaFilterModeLinear;
    texRef.normalized     = true;

    // Bind the array to the texture reference
    cudaBindTextureToArray(texRef, cuArray, channelDesc);

    // Allocate GPU buffers for the output image ..
    cudaMalloc(&d_output, sizeof(float) * width * height);

    dim3 threadsPerBlock(16,16);
    dim3 numBlocks((width/threadsPerBlock.x) + 1, (height/threadsPerBlock.y) + 1);

    Kernel<<<numBlocks, threadsPerBlock>>>(d_output, width,height);

    cudaDeviceSynchronize();

    // Copy output vector from GPU buffer to host memory.
    cudaMemcpy(h_output, d_output, sizeof(float) * width * height, cudaMemcpyDeviceToHost);

    // Free GPU  memory ...
}

コードで言ったように; このカーネルはテクスチャから読み取って、同じ画像を出力する必要があります。ただし、出力用に空の(灰色の)画像を使用しています。チュートリアルで同じ方法を実装したばかりですが、このテクスチャが機能しないのはなぜですか?

誰かがこの問題を解決する方法を教えてくれれば幸いです...

PS: 確かに、それはすべてのコードではありません。必要な部分だけコピペしました。他の詳細が必要な場合は、私もサポートします。

前もって感謝します。

4

1 に答える 1

6

正規化された座標を使用すると、テクスチャは 0 から 1 (排他的) の座標を介してアクセスされます。整数の threadIdx ベースの座標を正規化された座標に変換するのを忘れました。

unsigned int x = blockIdx.x * blockDim.x + threadIdx.x;
unsigned int y = blockIdx.y * blockDim.y + threadIdx.y; 
float u = x / (float)width;  
float v = y / (float)height;
于 2012-05-12T18:25:15.970 に答える