CUDAでdim3構造体の定数配列を定義する方法

Question

私はデバイス上で動作するようにいくつかのCUDAコードを書いています。このコードでは、定数値のルックアップテーブルを2つ使用します。これらの最初は、256個の符号なしintの配列であり、私はとしてそれを宣言する：

__constant__ 
uint16_t edgeTable[256]={ 
    0x000, 
    0x019, 
    ... etc. 
}; 

そして、これは罰金コンパイルするようです。

第二はDIM3の固定サイズアレイであり、私はこの試み：

__constant__ 
dim3 offsets[8] = { 
    {0, 0, 0}, {0, 0, 1}, {0, 1, 0}, 
    ... etc 
}; 

するためにどのコンパイラ・オブジェクト。エラーメッセージが表示されて：

error: dynamic initialization is not supported for __device__, __constant__ and __shared__ variables. 

はおそらく、私は、動的初期化を誤解したが、静的な初期化であるように私には思える、コンパイラは、すべてのサイズをうまくすることができ、すべての値が提供されています。

私はここで何が欠けていますか？

私は何をしようとしていますか？私はこの問題の重要な特徴は、それはCUDAはC++のコンパイルモデルを使用し、dim3は次のように扱われているのgcc 4.8.4

Answer 1

でのUbuntu 14.04にCUDA7.5ツールキットを使用してい

おかげ

クラス。そうしながら：それは一定のメモリ・オブジェクトを動的に初期化することを意味するので

__constant__ dim3 foo = {1,1,1}; 

はなく、：

dim3 foo = {1,1,1}; 

この理由は、パラメータ化コンストラクタの初期化のサポート、C++ 11に合法ですCUDA実行モデルでは許可されていません。

定数メモリの局面は、あなたにとって重要であり、あなたがdim3の利便性をしたい場合は、このような何か行うことができます：

#include <cstdio> 

__constant__ int offsets[3*8]; 

__global__ void kernel() 
{ 
    if (threadIdx.x < 8) { 
     dim3 val = *reinterpret_cast<dim3*>(&offsets[3*threadIdx.x]); 
     printf("%d (%d,%d,%d)\n", threadIdx.x, val.x, val.y, val.z); 
    } 
} 

void setup_offsets() 
{ 
    // This requires C++11 support 
    dim3 ovals[8] = { {0,0,0}, 
         {1,0,0}, {0,1,0}, {0,0,1}, 
         {1,1,0}, {1,0,1}, {0,1,1}, 
         {1,1,1} }; 

    cudaMemcpyToSymbol(offsets, &ovals[0], sizeof(ovals)); 
} 

int main(void) 
{ 
    setup_offsets(); 
    kernel<<<1,8>>>(); 
    cudaDeviceSynchronize(); 
    cudaDeviceReset(); 
    return 0; 
} 

ビットハックですが、おそらくあなたは、下のために最善を期待することができます状況。そのコードのPTXを見ると、コンパイラはの各メンバーをフェッチするためにld.const.u32を正しく出力しました。