cuda - 乘以 3D 矩阵和 2D 矩阵 CUDA

Question

3D 矩阵中的数据是由层(从上到下)生成的，我想将该数据与 2D 矩阵相乘 B 但我需要取一个向量而不是获取每一层来自第 1 层的向量，来自第 2 层的向量等等。

目前我正在做的是将这些向量从 3D 矩阵复制到 2D 矩阵 tmpA 然后与 B 相乘(使用 CUBLAS)并将结果存储在 < strong>tmpB 最终逐行复制回它在 3D 矩阵 C 中对应的位置。

总的来说，我的整个应用程序的运行速度至少比 CPU 版本快两倍，但在我看来，那些从设备到设备的内存副本(甚至)对于性能而言并不是很好。

执行此计算的更好方法是什么？我正在考虑在相乘之前重新排列数据，以避免内存复制。

3D 矩阵 A 和 C 以及 2D 矩阵 B 已经在 GPU 的内存中。

编辑

令 M、N、P 为 3D 矩阵 A 的维度，该矩阵以行主顺序存储在设备内存中的线性阵列中。我的代码如下所示:

cudaMalloc((void**)&d_tmpIn, sizeof(float)*M*P);
cudaMalloc((void**)&d_tmpOut, sizeof(float)*M*P);
cudaMalloc((void**)&d_C, sizeof(float)*M*N*P);

for (int iN = 0; iN < N; iN++)
{
  dst = d_tmpIn;
  for (int iM = 0; iM < M; iM++)
  {
    cudaMemcpy(dst, &(d_A[iN*P+0+iM*N*P]), sizeof(float)*P, cudaMemcpyD2D);
    dst += P;
  }

  cublasDgemm(cublasHandle, CUBLAS_OP_N, CUBLAS_OP_N, P, M, M, &alpha, d_tmpIn, P, d_B, M, &beta, d_tmpOut, P);

  src = d_tmpOut;
  for (int iM = 0; iM < M; iM++)
  {
    cudaMemcpy(&(d_C[iN*P+0+iM*N*P]), src, sizeof(float)*P, cudaMemcpyD2D);
    src += P;
  }
}

希望这对您有所帮助。

Answer 1

你不需要做内存拷贝! BLAS 和 LAPACK API 的创建方式使您可以指定起点、步幅长度、前导尺寸的长度等。

这样您就可以按原样使用 3D 数组 A 和 C，但使用正确的参数调用 cublasDgemm。

在您的情况下(如果我正确理解代码)，看起来每个矩阵都应该是 P X M 并且您有 N 个矩阵。但看起来 3D 数组排列为 PxNxM。因此，如果不为 d_tmpIn 和 d_tmpOut 分配内存，您可以这样做: A 的行数是 P。列数是 M。但是，前导维度 (lda) 应表示为 N * P。 C 也是如此。

int lda = N * P;
int ldc = N * P;
for (int iN = 0; iN < N; iN++)
{
  double *d_tmpIn = d_A + iN * P;
  double *d_tmpOut = d_C + iN * P;
  cublasSetStream(streams[iN]); // Optional
  cublasDgemm(cublasHandle, CUBLAS_OP_N, CUBLAS_OP_N,
              P, M, M, &alpha, d_tmpIn, lda, d_B, M, &beta, d_tmpOut, ldc);

}

您还可以创建 iN 流并在单独的流中运行每个 cublas。请注意，这仅在 M 和 P 足够小(即 GPU 在计算上尚未饱和)时才有用

编辑 如果您打算继续使用流，请尝试在程序开始时创建一次并重新使用它们。不要在与 Dgemm 相同的循环中创建和销毁流。这会增加开销。