如何在CUDA中解決「預期的標識符」錯誤

Question

編譯第5行時，我在CUDA C編程中遇到內核問題。我得到了"expected an identifier"錯誤。這是爲什麼發生？

我的內核函數如下：

__global__ void txz_kernel(float *txz,float *vz) 
{ 
int x = threadIdx.x + blockIdx.x * blockDim.x; 
int y = threadIdx.y + blockIdx.y * blockDim.y; 
int offset = x + y * blockDim.x * gridDim.x; --> error 
txz[offset]=txz[offset]+vz[offset]; 
} 

這裏是全碼：

#include "../common/book.h" 
#include "conio.h" 
#include "cuda.h" 
#include <fstream> 
#include <sstream> 
#include <iostream> 
#include <assert.h> 
#include "../common/book.h" 
#include <fstream> 
#define DIMX 320 
#define DIMZ 320 
#define PI 3.1415926535897932f 

__global__ void txz_kernel(float *txz,float *vz) 
{ 
int x = threadIdx.x + blockIdx.x * blockDim.x; 
int y = threadIdx.y + blockIdx.y * blockDim.y; 
int offset = x + y * blockDim.x * gridDim.x; --> error 
txz[offset]=txz[offset]+vz[offset]; 
} 

int main(void) { 
    float    *txz; 
    float    *vz; 

     HANDLE_ERROR(cudaMalloc((void**)&txz, DIMX * DIMZ * sizeof(float))); 
     HANDLE_ERROR(cudaMalloc((void**)&vz, DIMX * DIMZ * sizeof(float))); 

     float *tempvz = (float*)malloc(sizeof(float)*(DIMX*DIMZ));  
     float *temptxz = (float*)malloc(sizeof(float)*(DIMX*DIMZ)); 

    for (int i=0; i<DIMX; i++) { 
     for (int j=0; j<DIMZ; j++) { 
     int ij=DIMX*j + i; 
     tempvz[ij]=200.0; 
     temptxz[ij]=100.0; 
     } 
    } 

    for (int i=0; i<DIMX; i++) { 
     for (int j=(121); j<DIMZ; j++) { 
     int ij=DIMX*j + i;  
     tempvz[ij]=250.0; 
     temptxz[ij]=150.0; 
     } 
    } 
      HANDLE_ERROR(cudaMemcpy(vz, tempvz,sizeof(float)*(DIMX*DIMZ),cudaMemcpyHostToDevice)); 
      HANDLE_ERROR(cudaMemcpy(txz, temptxz,sizeof(float)*(DIMX*DIMZ),cudaMemcpyHostToDevice)); 
       dim3 blocks(DIMX/16,DIMZ/16); 
       dim3 threads(16,16); 
      txz_kernel<<<blocks,threads>>>(txz,vz) ;    
} 

Answer 1

你必須把這個代碼放到一個文件與.CU文件擴展名，並用NVCC編譯。 nvcc使用文件擴展名來確定任何給定輸入文件的編譯軌跡，如果該文件沒有.cu文件擴展名，則假定輸入中沒有設備代碼並將其傳遞給主機編譯器。代碼沒有問題，只是你沒有正確編譯它。

讓我們先從你的內核代碼在.cpp文件：

> type txzkernel.cpp 
__global__ void txz_kernel(float *txz, float *vz) 
{ 
    int x = threadIdx.x + blockIdx.x * blockDim.x; 
    int y = threadIdx.y + blockIdx.y * blockDim.y; 
    int offset = x + y * blockDim.x * gridDim.x; 
    txz[offset]=txz[offset]+vz[offset]; 
} 

現在讓我們嘗試用NVCC編譯它：

> nvcc -arch=sm_20 -Xptxas="-v" -c txzkernel.cpp 
txzkernel.cpp 
txzkernel.cpp(1) : error C2144: syntax error : 'void' should be preceded by ';' 
txzkernel.cpp(1) : error C4430: missing type specifier - int assumed. Note: C++ 
does not support default-int 
txzkernel.cpp(3) : error C2065: 'threadIdx' : undeclared identifier 
txzkernel.cpp(3) : error C2228: left of '.x' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(3) : error C2065: 'blockIdx' : undeclared identifier 
txzkernel.cpp(3) : error C2228: left of '.x' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(3) : error C2065: 'blockDim' : undeclared identifier 
txzkernel.cpp(3) : error C2228: left of '.x' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(4) : error C2065: 'threadIdx' : undeclared identifier 
txzkernel.cpp(4) : error C2228: left of '.y' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(4) : error C2065: 'blockIdx' : undeclared identifier 
txzkernel.cpp(4) : error C2228: left of '.y' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(4) : error C2065: 'blockDim' : undeclared identifier 
txzkernel.cpp(4) : error C2228: left of '.y' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(5) : error C2065: 'blockDim' : undeclared identifier 
txzkernel.cpp(5) : error C2228: left of '.x' must have class/struct/union 
     type is ''unknown-type'' 
txzkernel.cpp(5) : error C2065: 'gridDim' : undeclared identifier 
txzkernel.cpp(5) : error C2228: left of '.x' must have class/struct/union 
     type is ''unknown-type'' 

結果是很多語法錯誤的，因爲主機編譯器（在這種情況下，Microsoft Visual C++）不理解內核代碼中的任何CUDA語言擴展。

現在重命名文件，使其具有正確的擴展和重新編譯：

> rename txzkernel.cpp txzkernel.cu 
> nvcc -arch=sm_20 -Xptxas="-v" -c txzkernel.cu 
txzkernel.cu 
tmpxft_000012dc_00000000-3_txzkernel.cudafe1.gpu 
tmpxft_000012dc_00000000-8_txzkernel.cudafe2.gpu 
txzkernel.cu 
ptxas info : Compiling entry function '_Z10txz_kernelPfS_' for 'sm_20' 
ptxas info : Function properties for _Z10txz_kernelPfS_ 
    0 bytes stack frame, 0 bytes spill stores, 0 bytes spill loads 
ptxas info : Used 5 registers, 40 bytes cmem[0] 
tmpxft_000012dc_00000000-3_txzkernel.cudafe1.cpp 
tmpxft_000012dc_00000000-14_txzkernel.ii 

沒有錯誤。如果使用nvcc編譯內核代碼，並將內核代碼放入具有適當擴展名的文件中，則此代碼將不加修改地編譯。