cuda & cublas :使用cublas后调用全局函数

Question

我编写了一个程序，涉及矩阵向量乘法和最小二乘的所有使用cublas & cula。这个程序将重复多次。在每一步中，我必须将一个矩阵的特定行全部设置为零。

我试图将整个矩阵(50*1000或更高)复制到cpu中，并将一行设置为零，然后将矩阵复制回来，但这太费时了，因为程序会迭代10次或更多次。所以我决定写一个内核函数。

像这样的全局函数：

__global__ void Setzero(float* A, int index) /* A is the matrix and in col-major , index is the row I want to set zero */
{
    int ind=blockDim.x*blockIdx.x+threadIdx.x;
    if( ((ind%N)==index ) && (ind<50000) )  //notice matrix is in col-major ,matrix size is 50000
    {   
    A[ind]=0.0;
        ind+=blockDim.x*blockIdx.x;
    }
    else    ;
        __syncthreads();   
}

问题是当我这样做时(在调用函数之前使用cublas )：

cudaMalloc((void**)&A_Gpu_trans,sizeof(float)*50000);
cudaMemcpy(A_Gpu_trans,A_trans,sizeof(float)*M*N,cudaMemcpyHostToDevice);
cublasSgemv_v2(handle,CUBLAS_OP_N,1000,50,&al,A_Gpu_trans,1000,err_gpu,1,&beta,product,1);
dim3 dimBlock(16,1);
dim3 dimGrid((50000-1)/16+1,1);
Setzero<<<dimGrid,dimBlock>>>(A_Gpu_trans,Index);

它返回错误：

a __host__ function("Setzero") redeclared with __global__.

还有另一个错误：

MSB3721:命令“C：\Program\NVIDIA计算Toolkit\CUDA\v5.5\bin\nvcc.exe”-gencode=arch=compute_10，code=\"sm_10，compute_10\“-使用-本地-env-cl-版本2010 -ccbin "D:\Program \Microsoft 10.0\VC\bin”-I"C:\Program \NVIDIA计算工具包\CUDA\v5.5\包括“-I"C:\Program \NVIDIA计算工具包\cl\v5.5\包括”-G -保持-dir调试-maxrregcount=0 -机器32 -编译-cudart静态-g -DWIN32 -D_DEBUG -maxrregcount=0“-D_CONSOLE -D_MBCS -Xcompiler "/EHsc /W3 /nologo /Od /Zi /RTC1 /MDd“-o Debug\kernel.cu.obj "C:\Users\Administrator\documents\visual studio 2010\Projects\OOmp\OOmp\kernel.cu”“返回2。

奇怪的是，当我只使用cublas & cula时，我可以得到正确的答案。

Answer 1

虽然您没有在您的问题中显示它，但显然在代码中有另一个名为Setzero的主机函数。简单的解决方案是将内核重命名为其他内容。

CUDA工具链发出错误的根本原因是运行时API中的Setzero<<< >>>内核调用语法导致CUDA前端创建与内核同名的主机函数和匹配的参数列表，并将内核启动替换对该函数的调用。这个主机函数包含启动内核所需的API调用。通过使用与内核同名的另一个主机函数，可以避免此过程，并导致所看到的编译错误。

Answer 2

而且，你的功能是错误的，而且效率很低.

您不能在这样的条件中使用同步线程调用，它可能会导致挂起。这里似乎也完全没有必要。

更重要的是，您正在为每个矩阵条目启动一个线程，其中只有1/N实际上做了任何事情。

更好的方法是只启动与条目相对应的线程，这些线程将被设置为零。就像这样：

__global__ void Setzero(float* A, int index) 
{
  int ind=blockDim.x*blockIdx.x+threadIdx.x;
  if (ind < M)   
    A[index+N*ind]=0.0;
}

然后启动M线程(或者更确切地说，ceil(M/ 256 )线程块，每个线程由256个线程组成，或者任意块大小)。

例如：

int block_size = 256; // usually a good choice
int num_blocks = (M + block_size - 1) / block_size;
Setzero<<<num_blocks, block_size>>>(A, index);