编写CUDA代码的低加速比

Question

我用CPU (C++)和CUDA为同一个程序写了两个不同的代码。我不知道为什么CUDA代码的加速比CPU代码要小。

我有三个矩阵H，E，F，并对它们执行操作。CPU代码的加速时间为0.004s，CUDA代码的加速时间为0.006s，其中矩阵的维数为32*32。在内核代码中，我定义了三个共享内存变量matrix_H，matrix_E，matrix_F，并将dev_H，dev_E，dev_F值从全局内存复制到共享内存，以加快内存的访问时间，最后将计算出的共享内存变量复制到全局内存。

是因为内核调用或其他地方有很多参数吗？

__global__ void kernel_ScoreMatrix(char *dev_seqA, char *dev_seqB, 
   int *dev_H, int *dev_E, int *dev_F, int *dev_i_side, int *dev_j_side,
   int *dev_lenA, int *dev_idx_array, int *dev_array_length)
{
   __shared__ int matrix_H[1024];
   __shared__ int matrix_E[1024];
   __shared__ int matrix_F[1024];

   int x= threadIdx.x;
   int y= threadIdx.y;

   //calculate current_cell that execute with threads
   int current_cell = *(dev_lenA)*(y) + x;

   matrix_H[current_cell]=dev_H[current_cell];
   matrix_E[current_cell]=dev_E[current_cell];
   matrix_F[current_cell]=dev_F[current_cell];

   int index=0;

   int scoreMatrix[4];

   //for determine cells  that must compute in this time
   for (int i=0; i<*(dev_array_length); i++)
    if (current_cell== dev_idx_array[i]){
            scoreMatrix[0] = H_Matrix(current_cell, x, y, matrix_H, dev_seqA, dev_seqB, dev_lenA); 
            scoreMatrix[1] = E_Matrix(current_cell, matrix_E, matrix_H, dev_lenA);
            scoreMatrix[2] = F_Matrix(current_cell, matrix_F, matrix_H, dev_lenA);
            scoreMatrix[3] = 0;
            dev_H[current_cell] = findMax(scoreMatrix,4, index);
}

在main函数中：

dim3 threadsPerBlock(32, 32);
kernel_ScoreMatrix<<<1,threadsPerBlock>>>(dev_seqA, dev_seqB, dev_H, dev_E, dev_F, 
        dev_i_side, dev_j_side, dev_lenA, dev_idx_array, dev_array_length);

Answer 1

线程块by definition在单个SM上执行。因此，无论该线程块包含多少个线程，唯一可用于执行该特定线程块的执行资源是该(单个) SM中的资源。由于几乎所有NVIDIA GPU都包含多个SM，为了让GPU保持繁忙(这是获得最高性能所必需的)，因此有必要启动具有多个线程块的网格。一个合理的经验法则是，线程块的数量至少是SMs的2-4倍，通常情况下，拥有更多的线程块没有什么坏处。

但是如果你启动一个只有1个线程块的内核，你就会被限制为1个SM。因此，您可以从计算机获得大约1/(number of SMs in your GPU)的可用性能。该线程块中的线程数不会影响此系数。