2012-01-23 14 views
2

OpenCLにバッファとして複雑なデータ型を渡す必要があり、可能であればバッファの整列を避けたいと思います。OpenCLでのデータ整列の回避

私は彼らにキャストバッファに渡されたデータを区別するために2つの構造体を使用する必要があるのOpenCLでは

typedef struct 
{ 
    char a; 
    float2 position; 
} s1; 

typedef struct 
{ 
    char a; 
    float2 position; 
    char b; 
} s2; 

私はこのようにカーネルを定義:

__kernel void 
Foo(
    __global const void* bufferData, 
    const int amountElements // in the buffer 
) 
{ 
    // Now I cast to one of the structs depending on an extra value 
    __global s1* x = (__global s1*)bufferData; 

} 

をそして、それがうまく機能バッファに渡されたデータを整列させるときだけです。

質問です:使用する方法はあり_に渡されたデータで属性 _((パック))または_ 属性 _((整列(1)))を回避するために、アライメントバッファー?

答えて

1

小さい構造をパディングすることはできない場合は、私はあなたのカーネル関数は、タイプが何であるかを知っているように、別のパラメータを渡すことをお勧め - 要素の多分ちょうどサイズ。

データタイプが9バイトと10バイトであるため、カーネル内で読み取ったデータの数に応じて、両方を12バイトにパッディングしてみる価値があります。

あなたが興味があるかもしれない何か他のものを拡張したものです:cl_khr_byte_addressable_store http://www.khronos.org/registry/cl/sdk/1.0/docs/man/xhtml/cl_khr_byte_addressable_store.html

アップデート:私はあなたが混在配列を集めて実現しなかった 、私はそれはタイプに均一であったと思いました。要素ごとにタイプをトラッキングする場合は、タイプ(またはコード)のリストを渡す必要があります。 bufferDataでそれ単独でfloat2を使用する方が、おそらくより高速になります。

__kernel void 
Foo(
    __global const float2* bufferData, 
    __global const char* bufferTypes, 
    const int amountElements // in the buffer 
) 
+0

こんにちは@ mfa、あなたはパディングはオプションではありませんか?私が実際にバッファーに渡す必要があるのは、2つの異なるタイプのノードを持つツリーなので、作業は少し難しいでしょう。私はそれがどのタイプのノードであるかを示す各ノードに1バイトを渡しています。 – Michelle

+0

あなたは正確に何をしようとしていますか?私は別の提案で上記の更新を行いました。 – mfa

+0

私は、GPUメモリにツリーを送信しようとしているが、パディングがある場合、私は、私は、バッファの設計とOpenCLの中のプログラムでのケアの多くを取る必要がありますので、メモリのパディングを避けるためにしようとしたがよ私はそれをしなければならないでしょう。 パディングを行うアルゴリズムを作成することを考えていました。私は** float3 **を渡さなければなりません。** float2 **として渡すと、OpenCLのコードが不明瞭になりますので、埋め込みを行い、明確なOpenCLのコードを持つことを好みます。 ご協力いただきありがとうございます – Michelle

関連する問題