在并行计算中转置不同的矩阵。

Question

在并行计算中转置不同的矩阵。

3

我有三个不同大小的矩阵，想要并行转置它们。首先，我使用malloc将它们放入一个二维数组中，然后使用cudaMalloc将数组从主机（h_B）传输到设备（d_B）。使用threadIdx找到数组中每个矩阵的地址。使用cublas函数进行操作。以下是我的代码。

这段代码可以编译，但我无法得到结果。似乎在全局函数float *A = new float[m*n]中使用new不是一个好方法。请问有什么解决方法吗？非常感谢！

/* Includes, system */
#include <stdio.h>
#include <stdlib.h>
#include<iostream>

/* Includes, cuda */
#include <cuda_runtime.h>
#include <cublas_v2.h>

/* Includes, cuda helper functions */
#include <helper_cuda.h>

__global__ void transposeCublasSgeam(int *M_A, int *N_A, float *ptrA, float *ptrC, const int N, int *address)
{
    cublasHandle_t cnpHandle;
    cublasStatus_t status = cublasCreate(&cnpHandle);

    if (status != CUBLAS_STATUS_SUCCESS)
    {
        return;
    }

    const float d_alpha = 1.0f;
    const float d_beta = 0.0f;
    int idx = threadIdx.x;
    if(idx<N){
      int m = M_A[idx]; //A_row
      int n = N_A[idx]; //A_col
      float *A = new float[m*n]; 
      float *C = new float[m*n];
      A = ptrA+address[idx];
      C = ptrC+address[idx];
      cublasSgeam(cnpHandle, CUBLAS_OP_T, CUBLAS_OP_T, m, n, &d_alpha, (const float*)A, n, &d_beta, (const float *)A, n, C, m);
      delete[] A;
      delete[] C;
    }
     cublasDestroy(cnpHandle);

}

int main()
{

    const int N = 3;
    int M_B[N] = { 2,3,2 }; //row number of matrices 
    int N_B[N] = { 3,2,4 }; //col number of matrices 

    float a[6] = { 1,2,3,
                  4,5,6 };
    float b[6] = { 1,2,
                 3,4,
                 5,6};
    float c[8] = { 1,2,3,1,
                  2,3,4,5 };

    float **h_B = (float**)malloc(N * sizeof(float*));
    float **h_BT = (float**)malloc(N * sizeof(float*));

    h_B[0] = a, h_BT[0] = a;
    h_B[1] = b, h_BT[1] = b;
    h_B[2] = c, h_BT[2] = c;

    int NUM_B = 20; // total number of elements
    int address[] = {0,6,12}; 

    float *d_B, *d_BT;
    checkCudaErrors(cudaMalloc((void **)&d_B, NUM_B * sizeof(float)));
    checkCudaErrors(cudaMalloc((void **)&d_BT, NUM_B * sizeof(float)));
    checkCudaErrors(cudaMemcpy(d_B, h_B, NUM_B * sizeof(float), cudaMemcpyHostToDevice));
    checkCudaErrors(cudaMemcpy(d_BT, h_BT, NUM_B * sizeof(float), cudaMemcpyHostToDevice));

    transposeCublasSgeam<<<1,N>>>(M_B, N_B, d_B,d_BT, N,address); 

    checkCudaErrors(cudaMemcpy(h_BT, d_BT, NUM_B * sizeof(float), cudaMemcpyDeviceToHost));

    cudaFree(d_B);
    cudaFree(d_BT);
    delete[] h_B;
    delete[] h_BT;

    return 0;
}

- Mardha

你将 A 改为指向其他地方，然后调用 delete。即使在主机代码中，这也是一个错误。 - user9400869

cublas函数可以从设备中调用。我认为关键是如何从数组中获取矩阵的相应数据。在设备中分配A和C可能是错误的吗？ - Mardha

看起来你是对的，而且这是可能的：https://docs.nvidia.com/cuda/cuda-c-programming-guide/index.html#dynamic-global-memory-allocation-and-operations - user9400869

我怀疑A = prtA + address[x]这一行并不是你想要的。这会改变指针而不是值，从而无法访问已分配的内存。 - user9400869

它将指针更改为指向数组中的另一个矩阵。例如，第一个矩阵有6个元素，指针是ptrA [0]。对于矩阵的第二个数据，指针是ptrA + 6。 - Mardha

显示剩余4条评论

1个回答

网页内容由stack overflow 提供, 点击上面的

可以查看英文原文，
原文链接

- Robert Crovella · Accepted Answer

你的代码中有一些错误。在我的描述中可能会漏掉一些。

请注意，新版CUDA不再支持在设备代码中使用cublas功能。
每个传递给设备代码的指针都需要使用cudaMalloc进行分配。你已经为一些指针做了cudaMalloc，但并非全部。
你对指针和指针数组感到困惑。我无法为你解决所有问题。你的内核设计不需要使用指针数组的复杂性。因此，我已经将其全部删除。
在CUDA动态并行性（CDP）中，不能将指向本地地址空间的指针传递给子内核。你无法在CDP中使用alpha和beta在本地地址空间中，并将指向这些变量的指针传递给CUBLAS。
要进行纯转置，请查阅CUBLAS Sgeam文档以获取建议使用的参数。

我相信我还修复了其他问题。请参考这个示例：

$ cat t1433.cu
/* Includes, system */
#include <stdio.h>
#include <stdlib.h>
#include<iostream>

/* Includes, cuda */
#include <cuda_runtime.h>
#include <cublas_v2.h>

/* Includes, cuda helper functions */
#include <helper_cuda.h>

__global__ void transposeCublasSgeam(int *M_A, int *N_A, float *ptrA, float *ptrC, const int N, int *address)
{
    cublasHandle_t cnpHandle;
    cublasStatus_t status = cublasCreate(&cnpHandle);

    if (status != CUBLAS_STATUS_SUCCESS)
    {
        printf("thread: %d, error1: %d\n", threadIdx.x, (int)status);
        return;
    }

    float *d_alpha =  new float; // a pointer to device-heap, not local memory
    *d_alpha = 1.0f;
    float *d_beta  = new float;
    *d_beta = 0.0f;
    int idx = threadIdx.x;
    if(idx<N){
      int m = M_A[idx]; //A_row
      int n = N_A[idx]; //A_col
      status = cublasSgeam(cnpHandle, CUBLAS_OP_T, CUBLAS_OP_N, m, n, d_alpha, ptrA+address[idx], n, d_beta, ptrC+address[idx], m, ptrC+address[idx], m);
      if (status != CUBLAS_STATUS_SUCCESS)
      {
        printf("thread: %d, error2: %d\n", threadIdx.x, (int)status);
        return;
      }
    }
     cublasDestroy(cnpHandle);

}

int main()
{

    const int N = 3;
    int M_B[N] = { 2,3,2 }; //row number of matrices
    int N_B[N] = { 3,2,4 }; //col number of matrices

    float a[6] = { 1,2,3,
                  4,5,6 };
    float b[6] = { 1,2,
                 3,4,
                 5,6};
    float c[8] = { 1,2,3,1,
                  2,3,4,5 };
    float *h_Bdata  = (float *)malloc(sizeof(a)+sizeof(b)+sizeof(c));
    float *h_BTdata = (float *)malloc(sizeof(a)+sizeof(b)+sizeof(c));
    memcpy(h_Bdata, a, sizeof(a));
    memcpy(h_Bdata+(sizeof(a)/sizeof(a[0])), b, sizeof(b));
    memcpy(h_Bdata+(sizeof(a)/sizeof(a[0]))+(sizeof(b)/sizeof(b[0])), c, sizeof(c));

    int NUM_B = 20; // total number of elements
    int address[] = {0,6,12};
    int *d_address;
    cudaMalloc(&d_address, sizeof(address));
    cudaMemcpy(d_address, address, sizeof(address), cudaMemcpyHostToDevice);
    int *d_M_B, *d_N_B;
    cudaMalloc(&d_M_B, sizeof(M_B));
    cudaMalloc(&d_N_B, sizeof(N_B));
    cudaMemcpy(d_M_B, M_B, sizeof(M_B), cudaMemcpyHostToDevice);
    cudaMemcpy(d_N_B, N_B, sizeof(N_B), cudaMemcpyHostToDevice);
    float *d_B, *d_BT;
    checkCudaErrors(cudaMalloc((void **)&d_B, NUM_B * sizeof(float)));
    checkCudaErrors(cudaMalloc((void **)&d_BT, NUM_B * sizeof(float)));
    checkCudaErrors(cudaMemcpy(d_B, h_Bdata, NUM_B * sizeof(float), cudaMemcpyHostToDevice));

    transposeCublasSgeam<<<1,N>>>(d_M_B, d_N_B, d_B,d_BT, N,d_address);

    checkCudaErrors(cudaMemcpy(h_BTdata, d_BT, NUM_B * sizeof(float), cudaMemcpyDeviceToHost));
    std::cout << "B , BT" << std::endl;
    for (int i = 0; i < NUM_B; i++){
      std::cout << h_Bdata[i]  << " , " << h_BTdata[i] << std::endl;}
    cudaFree(d_B);
    cudaFree(d_BT);

    return 0;
}
$ /usr/local/cuda-8.0/bin/nvcc -I/usr/local/cuda-8.0/samples/common/inc  t1433.cu -rdc=true -lcublas_device -lcudadevrt -arch=sm_35 -o t1433
$ LD_LIBRARY_PATH=/usr/local/cuda-8.0/lib64 CUDA_VISIBLE_DEVICES="3" cuda-memcheck ./t1433
========= CUDA-MEMCHECK
B , BT
1 , 1
2 , 4
3 , 2
4 , 5
5 , 3
6 , 6
1 , 1
2 , 3
3 , 5
4 , 2
5 , 4
6 , 6
1 , 1
2 , 2
3 , 2
1 , 3
2 , 3
3 , 4
4 , 1
5 , 5
========= ERROR SUMMARY: 0 errors
$