2

次の定義を持つ構造体プリミティブがあります。

typedef struct Primitive {
    float m[12];
    float invm[12];
    enum PrimitiveType type;
    int rayDensity;
    float util1;             
    float util2;                
} Primitive;

これらの構造体の配列を、定数メモリバッファ内のカーネルに渡します。

__constant Primitive *objects;

最適化の一環として、構造体をローカルメモリにロードする方法を検討したいので、カーネルには次のようなコードがあります。

__kernel void test(int n_objects, __constant Primitives *objects) {
    local Primitive pFrom, pTo;

    for(int i = 0; i < n_objects; i++) {
        pFrom = objects[i]; 
    }

}

これを実行すると、次のようなコンパイルエラーが発生します。

ptxas application ptx input, line 42; error: State space mismatch between instruction and address in instruction 'ld'

実験として、最初に構造体をプライベート変数にコピーし、次にローカル変数に次のようにコピーしてみました。

__kernel void test(int n_objects, __constant Primitives *objects) {
    Primitive pF, Pt;
    local Primitive pFrom, pTo;

    for(int i = 0; i < n_objects; i++) {
        pF = objects[i]
        pFrom = pF; 
    }

}

これはコンパイルされて実行されますが、オブジェクトはローカル変数pFromに深くコピーされていないようです。

私のコードサンプルは純粋なサンプルであり、簡潔にするためにすべてを削除したことに注意してください。また、定数グローバルメモリから直接プリミティブ構造体を使用すると、コードは正常に機能します。

誰かが私がここで欠けているものを知っていますか、確かにそれはディープコピーまたはOpenCLアドレス空間のいくつかの基本的な基本です。

4

1 に答える 1

1

必要なのはasync_work_group_copy 関数です。この非同期操作が完了するまで待つには、wait_group_events関数を使用します。

お役に立てれば。

于 2012-06-18T14:06:15.363 に答える