1

opencl で変数 union(private) を使用すると、奇妙な動作が発生します。コード：

v = { 0, 1, 2, ... } // Defined in host and load to Device

typedef union svec8{
    int  word[8];
    int8 word8;
} vec8;

__kernel
void red( global vec8 *v, global int *out ){
    uint sizeBin = 8;
    vec8 binning = {0}; // Every Thread has a 8-space bin, initialized with 0
    uint gID     = get_global_id(0);
    int temp;

    binning.word8 = v[ dID ].word8;

    #ifdef CONDITION
        temp = 0;
        for ( uint i = 0; i < sizeBin; i++ ){
            temp += binning.word[ i ];
        }
    #endif

    if ( dID == 0 ){
        *out = binning.word[n]; // n belongs to [0, 7]
    }
}

問題は、a が選択するすべての n に対して、CONDITION が定義されている場合、*out は常に 1 に等しくなりますが、CONDITION を undef にすると、*out は正しい値、つまり、0、1、..、または 7 を取得することです。どちらを選択するかに応じてコース。

また、union の使用をやめて int8 だけを使用すると、問題が解決することにも気付きました。

前もって感謝します、

プラットフォーム: Ubuntu 12.04 - 3.2.0-24-generic-pae - ドライバー OpenCL 1.2 AMD-APP (923.1)

4

1 に答える 1

3

AMD にはユニオンの配置に関するバグがありますが、それが原因でしょうか? 私のレポートを参照してください。Intel SDK で同じコードを実行して、違いがあるかどうかを確認することをお勧めします。

編集:私自身のコード(多くの異なるユニオン)では、上流で修正されるまで、このソリューションを使用します:

#define AMD_UNION_ALIGN_BUG_WORKAROUND() __attribute__((aligned(32)))
//            this is the biggest alignment value in the union ^^^^^
// in your case sizeof(int)*8=32 (in bytes)

// define unions like this
union svec8 { /*...*/ } AMD_UNION_ALIGN_BUG_WORKAROUND() vec8;
于 2012-06-12T14:30:48.750 に答える