2012-03-14 5 views
3

いくつかのカーネル起動時に読みたいテクスチャがいくつかあります。ドキュメンテーションによれば、これらはグローバルに定義する必要があります。問題は、さまざまなテクスチャソースと同じカーネル関数(かなり複雑です)を呼び出したいということです。CUDAでのテクスチャリファレンスの柔軟な使用には回避策があります

texture<unsigned char, 2, cudaReadModeElementType> g_tex_a; 
texture<unsigned char, 2, cudaReadModeElementType> g_tex_b; 

__global__ void gpu_kernel_1() 
{ 
    // long complicated kernel 
    foo = tex2D(g_tex_a, x,y); 
} 

__global__ void gpu_kernel_2() 
{ 
    // long complicated kernel 
    bar = tex2D(g_tex_a, x,y); 
} 

main() 
{ 
    gpu_kernel_1<<<grid, block>>>(); 
    gpu_kernel_2<<<grid, block>>>(); 
} 

は、例えば、それはtex2Dとから読むべきテクスチャカーネルに渡すの方法を持ってすることはできませんg_tex_aまたはg_tex_b?コンパイラは、コンパイル時にテクスチャ参照を知る必要があるようです。 gpu_kernel_1とgpu_kernel_2のコードを別のテクスチャで再利用できるようにしたい。

テクスチャ参照オブジェクトがホストとデバイスコードによって利用される方法を実際に理解していません。現時点では、tex2D関数が異なるテクスチャリファレンスを使用するという唯一の変更を加えて、各カーネルのすべてのコードを複製するという恐ろしい解決策があります。 gpu_kernel_1_with_tex_a()、gpu_kernel_1_with_tex_b()。

他に解決策はありますか?ありがとう。

+1

カーネルに余分なパラメータを渡すことができず、 'if/switch'条件を使用できませんか?すべてのスレッドが同じパスになっていれば、相違はありません。テクスチャをパラメータとして渡すというあなたの本来の問題に関しては、わかりません。 – pQB

+1

@pQB:パラメータをテンプレートすることはさらに優れた解決策です。コンパイラのデッドコードを削除すると、カーネルの特定のインスタンスで使用されていないコードパスが削除されます。 – talonmies

+0

ありがとうございます。私は、コンパイラがカーネル関数のテンプレートを許可しないと予想していたので、テンプレートを試しませんでした。しかし私は確かにそのことを知らないと認めます。 – Robotbugs

答えて

1

テクスチャ参照を使用する前に、cudaBindTextureToArray()などのメモリにマップする必要があります。マッピングすると、操作するデータを決めることができます。

+0

OKです。私はそのバインディングを忘れていました。なぜなら、そのコードの一部がinit関数で一度設定されているからです。ありがとう! – Robotbugs

1

テクスチャ参照APIであるCUDA 5.0に加えて、テクスチャオブジェクトAPIがサポートされました。テクスチャ参照をグローバルに定義する代わりに、この新しいAPIを使用して、テクスチャオブジェクトをローカルで定義し、それを関数の引数としてカーネルに渡すことができます。 CUDAプログラミングガイド(3.2.10.1.1項)をご覧ください。

関連する問題