3

CUDA5.0を使用しています。コンパイラーを使用すると、カーネル内でホスト宣言されたint定数を使用できるようになることに気付きました。ただし、ホストで宣言されたfloat定数を使用するカーネルのコンパイルは拒否されます。この一見矛盾の理由を誰かが知っていますか？

たとえば、次のコードはそのままで問題なく実行されますが、カーネルの最後の行がコメント化されていない場合はコンパイルされません。

#include <cstdio>
#include <cuda_runtime.h>

static int   __constant__ DEV_INT_CONSTANT   = 1;
static float __constant__ DEV_FLOAT_CONSTANT = 2.0f;

static int   const        HST_INT_CONSTANT   = 3;
static float const        HST_FLOAT_CONSTANT = 4.0f;

__global__ void uselessKernel(float * val)
{
    *val = 0.0f;

    // Use device int and float constants
    *val += DEV_INT_CONSTANT;
    *val += DEV_FLOAT_CONSTANT;

    // Use host int and float constants
    *val += HST_INT_CONSTANT;
    //*val += HST_FLOAT_CONSTANT; // won't compile if uncommented
}

int main(void)
{
    float * d_val;
    cudaMalloc((void **)&d_val, sizeof(float));

    uselessKernel<<<1, 1>>>(d_val);

    cudaFree(d_val);
}
4

1 に答える 1

4

デバイスコードに定数番号を追加することは問題ありませんが、ホストメモリに保存されている番号をデバイスコードに追加することはできません。

コード内ののすべての参照は、その変数のaddrが参照されない場合、コンパイラー/オプティマイザーによってstatic const int値に置き換えることができます。3この場合、それはのよう#define HST_INT_CONSTANT 3であり、この変数にホストメモリは割り当てられません。

ただし、floatvarの場合、ホストメモリは常に割り当てられstatic const floatます。カーネルはホストメモリに直接アクセスできないため、コードはstatic const floatコンパイルされません。

C / C ++の場合、intよりも積極的に最適化できますfloat。

コメントがオンのときにコードが実行されるのは、CUDACIthinkのバグと見なすことができます。これstatic const intはホスト側のものであり、デバイスから直接アクセスできないようにする必要があります。

于 2013-02-21T18:24:41.110 に答える