1

C ++テンプレートを使用すると、NVIDIA NVCC（CUDA 4.0および4.1テスト済み）の奇妙な動作が見られます。私はそれを、振る舞いを示す簡単な例にまとめました。

これはすでにバグレポートの状態にあります。ただし、このサイトはバグや修正の信頼できるソースとして成長しているため、ここに接続します。だから、私はこのページを更新し続けます。

コード：

#include"stdio.h"

#define PETE_DEVICE __device__

template<class T, int N>  class ILattice;
template<class T>         class IScalar;
template<class T, int IL> struct AddILattice {};

template<class T>
PETE_DEVICE
void printType() {
  printf("%s\n",__PRETTY_FUNCTION__);
}

template<class T> class IScalar {
  T F;
};

template<class T, int N> class ILattice {
  T F[N];
};

template<class T, int N>
struct AddILattice<IScalar<T> , N> {
  typedef ILattice< T , N > Type_t;
};

#define IL 16

__global__ void kernel()
{
  printf("IL=%d\n",IL);  // Here IL==16

  typedef typename AddILattice<IScalar<float> ,IL>::Type_t Tnew;

  // This still works fine. Output:
  // void printType() [with T = ILattice<float, 16>]
  //
  printType<Tnew>();

  // Now problems begin: Output:
  // T=4 Tnew=0 IL=64
  // Here IL should still be 16
  // sizeof(Tnew) should be 16*sizeof(float)
  //
  printf("T=%d Tnew=%d IL=%d\n",sizeof(IScalar<float> ),sizeof(Tnew),IL);   
}   

int main()
{
    dim3  blocksPerGrid( 1 , 1 , 1 );
    dim3  threadsPerBlock( 1 , 1, 1);
    kernel<<< blocksPerGrid , threadsPerBlock , 48*1024 >>>( );

    cudaDeviceSynchronize();
    cudaError_t kernel_call = cudaGetLastError();
    printf("call: %s\n",cudaGetErrorString(kernel_call));

}

コンパイラがIL16から64に変わる理由はありますか？

4

1 に答える 1

2

間違った printf 変換を使用している可能性があります。%dは int を出力することを意味しますが、sizeof は int ではなく size_t を返します。さらに size_t 長さ修飾子を使用します (そして符号なしにします)。つまり、 に置き換え%dます%zu。

printf は (var-args リストのために) どの型が実際に渡されたかを知ることができないため、型変換は行われず、書式文字列による型のみを知ることができます。したがって、そこに正しいパラメータを渡す必要があります。size_t が int と同じサイズのシステムを使用している場合、コードは機能します (たとえば、多くの 32 ビット システム)。しかし、その事実に頼ることはできません。適切な変換を使用すると、そこに役立ちます。

（したがって、コンパイラが定数を変更するのではなく、間違って出力するだけです）

于 2012-05-03T09:54:19.173 に答える