cublasSdot 的工作速度比 cublasSgemm 慢

Question

在我的玩具示例中，我首先将大小为 32x32、100 000 的矩阵相乘，然后再次计算大小为 1024、100 000 的两个向量的标量积。第一个我使用cublasSgemm，第二个-cublasSdot。

因此，第一次计算的时间是530 msec，第二次计算的时间是10 000 msec。然而，为了将矩阵相乘，我们需要执行32^3 操作（乘加），而对于标量积只需执行1024=32^2 操作。

那么为什么我会得到这样的结果？代码如下：

__device__ float res;
void randomInit(float *data, int size)
{
    for (int i = 0; i < size; ++i)
        data[i] = rand() / (float)RAND_MAX;
}
int main(){
    cublasHandle_t handle;
    float out;
    cudaError_t cudaerr;
    cudaEvent_t start1, stop1,start2,stop2;
    cublasStatus_t stat;
    int size = 32;
    int num = 100000;

    float *h_A = new float[size*size];
    float *h_B = new float[size*size];
    float *h_C = new float[size*size];
    float *d_A, *d_B, *d_C;
    const float alpha = 1.0f;
    const float beta = 0.0f;
    randomInit(h_A, size*size);
    randomInit(h_B, size*size);
    cudaMalloc((void **)&d_A, size *size *sizeof(float));
    cudaMalloc((void **)&d_B, size *size * sizeof(float));
    cudaMalloc((void **)&d_C, size *size * sizeof(float));
    stat = cublasCreate(&handle);
    cudaEventCreate(&start1);
    cudaEventCreate(&stop1);
    cudaEventCreate(&start2);
    cudaEventCreate(&stop2);
    cublasSgemm(handle, CUBLAS_OP_N, CUBLAS_OP_N, size, size, size, &alpha, d_A, size, 
                d_B, size, &beta, d_C, size);
    cudaEventRecord(start1, NULL);
    cudaMemcpy(d_A, h_A, size *size * sizeof(float), cudaMemcpyHostToDevice);
    cudaMemcpy(d_B, h_B, size *size * sizeof(float), cudaMemcpyHostToDevice);
    for (int i = 0; i < num; i++){
        cublasSgemm(handle, CUBLAS_OP_N, CUBLAS_OP_N, size, size, size, &alpha, d_A, 
                        size, d_B, size, &beta, d_C, size);
    }
    cudaMemcpy(h_C, d_C, size*size*sizeof(float), cudaMemcpyDeviceToHost);
    cudaEventRecord(stop1, NULL);
    cudaEventSynchronize(stop1);
    float msecTotal1 = 0.0f;
    cudaEventElapsedTime(&msecTotal1, start1, stop1);
    std::cout <<"total time for MAtMul:" << msecTotal1 << "\n";
    cudaEventRecord(start2, NULL);
    cudaMemcpy(d_A, h_A, size *size * sizeof(float), cudaMemcpyHostToDevice);
    cudaMemcpy(d_B, h_B, size *size * sizeof(float), cudaMemcpyHostToDevice);
    for (int i = 0; i < num; i++){
        cublasSdot(handle, 1024, d_A , 1, d_B , 1, &res);
    }
    cudaEventRecord(stop2, NULL);
    cudaEventSynchronize(stop2);
    float msecTotal2 = 0.0f;
    cudaEventElapsedTime(&msecTotal2, start2, stop2);
    std::cout << "total time for dotVec:" << msecTotal2 << "\n";
    cublasDestroy(handle);
    cudaFree(d_A);
    cudaFree(d_B);
    cudaFree(d_C);
    delete[] h_A;
    delete[] h_B;
    delete[] h_C;
    return 1;
}

更新：我还尝试通过将向量视为1 by 1024 矩阵来执行与cublasSgemm 的点积。结果是3550 msec，这更好，但仍然是第一次计算的7倍。

Answer 1

一个问题是您没有正确处理指针模式以调用cublasSdot。

您需要阅读手册中的this section。

还有这个：

    cublasSdot(handle, 1024, d_A , 1, d_B , 1, &res);
                                               ^^^^

在任何情况下都是非法的。在 CUDA 中，在主机代码中获取设备变量的地址是不合法的。你当然可以，但结果很垃圾。

当我如下修改你的代码时：

cublasSetPointerMode(handle, CUBLAS_POINTER_MODE_DEVICE);
float *dres;
cudaMalloc(&dres, sizeof(float));
cudaEventRecord(start2, NULL);
cudaMemcpy(d_A, h_A, size *size * sizeof(float), cudaMemcpyHostToDevice);
cudaMemcpy(d_B, h_B, size *size * sizeof(float), cudaMemcpyHostToDevice);
for (int i = 0; i < num; i++){
    if(cublasSdot(handle, 1024, d_A , 1, d_B , 1, dres) != CUBLAS_STATUS_SUCCESS) {std::cout << ".";}
}

cublasSdot 与 cublasSgemm 的执行时间比率约为 2:1，这可能是合理的，尤其是对于这些大小。在引擎盖下，点操作意味着并行减少。 1024 个线程可以计算部分结果，但随后需要 1024 个线程范围的并行缩减。 gemm 不需要并行减少，因此可能更快。可以分配 1024 个线程以在单个线程中生成 1024 个结果。对于内存绑定算法，32^2 和 32^3 操作之间的差异可能不是那么显着，但并行减少意味着显着的额外操作。然后，当我将程序中的 size 从 32 更改为 128 时，我看到比率反转，并且矩阵乘法确实比点积长 3 倍。