CUDAとNVCC：floatまたはdouble

Question

の間で問題選択するプリプロセッサを使用して：.Hを持つ

を、私はC/C++用や計算能力を持つCUDA用にコンパイル場合は2倍に本物を定義したいです> = 1.3。コンピューティング能力が<のcudaのためにコンパイルする場合は、realをfloatに定義します。私はこれに来た多くの時間（動作しない）

 
# if defined(__CUDACC__) 

#  warning * making definitions for cuda 

#  if defined(__CUDA_ARCH__) 
#   warning __CUDA_ARCH__ is defined 
#  else 
#   warning __CUDA_ARCH__ is NOT defined 
#  endif 

#  if (__CUDA_ARCH__ >= 130) 
#      define real double 
#      warning using double in cuda 
#  elif (__CUDA_ARCH__ >= 0) 
#    define real float 
#    warning using float in cuda 
#    warning how the hell is this printed when __CUDA_ARCH__ is not defined? 
#  else 
#    define real 
#    error what the hell is the value of __CUDA_ARCH__ and how can I print it 
#  endif 

# else 
#  warning * making definitions for c/c++ 
#  define real double 
#  warning using double for c/c++ 
# endif 

私は（-archフラグに注意してください）

 
nvcc -arch compute_13 -Ilibcutil testFloatDouble.cu 

をコンパイルするとき、私は

 
* making definitions for cuda 
__CUDA_ARCH__ is defined 
using double in cuda 

* making definitions for cuda 
warning __CUDA_ARCH__ is NOT defined 
warning using float in cuda 
how the hell is this printed if __CUDA_ARCH__ is not defined now? 

Undefined symbols for architecture i386: 
    "myKernel(float*, int)", referenced from: .... 

を取得した後

私はファイルがnvccによって2回コンパイルされることを知っています。最初のものはOKです（CUDACCが定義され、CUDA_ARCH> = 130）が、2回目はどうなりますか？ CUDA_DEFINED but CUDA_ARCH未定義または値< 130？どうして ？

お時間をいただきありがとうございます。今の私が見る唯一の現実的な解決策は、カスタムを使用している

Answer 1

を行います。両者には微妙な違いがあります。 __CUDA_ARCH__が最初の質問に回答し、__CUDACC__が2番目の質問に答えます。我々はCUDAアーキテクチャ依存のインスタンス化、nvccによってsteeeredホスト・コード用に別のスタンザ、およびホスト・コードのコンパイルのためのスタンザを持つテンプレートCUDAカーネルが操縦していない。ここ

#ifdef __CUDACC__ 
#warning using nvcc 

template <typename T> 
__global__ void add(T *x, T *y, T *z) 
{ 
    int idx = threadIdx.x + blockDim.x * blockIdx.x; 

    z[idx] = x[idx] + y[idx]; 
} 

#ifdef __CUDA_ARCH__ 
#warning device code trajectory 
#if __CUDA_ARCH__ > 120 
#warning compiling with double precision 
template void add<double>(double *, double *, double *); 
#else 
#warning compiling with single precision 
template void add<float>(float *, float *, float *); 
#else 
#warning nvcc host code trajectory 
#endif 
#else 
#warning non-nvcc code trajectory 
#endif 

：

は、次のコードスニペットを考えてみましょうby nvcc。あるここからポイントを奪う

$ ln -s cudaarch.cu cudaarch.cc 
$ gcc -c cudaarch.cc -o cudaarch.o 
cudaarch.cc:26:2: warning: #warning non-nvcc code trajectory 

$ nvcc -arch=sm_11 -Xptxas="-v" -c cudaarch.cu -o cudaarch.cu.o 
cudaarch.cu:3:2: warning: #warning using nvcc 
cudaarch.cu:14:2: warning: #warning device code trajectory 
cudaarch.cu:19:2: warning: #warning compiling with single precision 
cudaarch.cu:3:2: warning: #warning using nvcc 
cudaarch.cu:23:2: warning: #warning nvcc host code trajectory 
ptxas info : Compiling entry function '_Z3addIfEvPT_S1_S1_' for 'sm_11' 
ptxas info : Used 4 registers, 12+16 bytes smem 

$ nvcc -arch=sm_20 -Xptxas="-v" -c cudaarch.cu -o cudaarch.cu.o 
cudaarch.cu:3:2: warning: #warning using nvcc 
cudaarch.cu:14:2: warning: #warning device code trajectory 
cudaarch.cu:16:2: warning: #warning compiling with double precision 
cudaarch.cu:3:2: warning: #warning using nvcc 
cudaarch.cu:23:2: warning: #warning nvcc host code trajectory 
ptxas info : Compiling entry function '_Z3addIdEvPT_S1_S1_' for 'sm_20' 
ptxas info : Used 8 registers, 44 bytes cmem[0] 

：これは以下のように振る舞う

• __CUDACC__はnvccは、コンパイルを操舵か
• __CUDA_ARCH__されているかどうかを定義し、常に未定義ホスト・コードをコンパイルされ、操縦nvccまたは
• __CUDA_ARCH__は、compilaのデバイスコードの軌跡に対してのみ定義されていますションは、情報のこれらの3つの作品は、常に異なるCUDAアーキテクチャ、ホスト側のCUDAコード、およびコード全くnvccでコンパイルされていないと、デバイスコードの条件付きコンパイルを持っているのに十分ですnvcc

によって操縦します。 nvccのドキュメントは時々少し簡潔ですが、これはすべてコンパイルの軌道に関する議論でカバーされています。

Answer 2

は定義：

 

# if (!defined(__CUDACC__) || defined(USE_DOUBLE_IN_CUDA)) 
#  define real double 
#  warning defining double for cuda or c/c++ 
# else 
#  define real float 
#  warning defining float for cuda 
# endif 

、その後

 
nvcc -DUSE_DOUBLE_IN_CUDA -arch compute_13 -Ilibcutil testFloatDouble.cu 

、2つのコンパイルのために出力したよう：

 
#warning defining double for cuda or c/c++ 
#warning defining double for cuda or c/c++ 

および

NVCCはCUDAコードを処理している場合に、ホストとデバイスのコンパイル軌跡を区別するか、そしてどのようにCUDAと非CUDAコードを区別するために -

は、あなたが二つのことをconflatingかもしれないようだ

 
#warning defining float for cuda 
#warning defining float for cuda