홈>2020-02-01 00:07

nx1 벡터에서 mxn 행렬의 Cuda 곱셈

다음 커널은 두 개의 nxn 행렬을 곱합니다 :

   __global__ void matrixMultiplication(const double *A, const double *B, double *C, int N)
{
    int i = blockDim.y * blockIdx.y + threadIdx.y;
    int j = blockDim.x * blockIdx.x + threadIdx.x;
    double value = 0;
    for(int k = 0; k < N; k++){
    value += A[k * N + j] * B[i * N + k];
    }
    C[i * N + j] = value;
    }

MATLAB에서 위의 커널을 다음과 같이 사용합니다 :

k = parallel.gpu.CUDAKernel('matrixMultiplication.ptx', 'matrixMultiplication.cu');
A = rand(3,4);
b = rand(4,1);
C = zeros(3,1);
k.ThreadBlockSize = [3,4,1];
k.GridSize = [1, 1];
D = A*b;
C = feval(k,A,b,C,4);
D-C

그러나 결과는 0이 아닙니다! nx1 벡터에서 mxn 행렬을 곱할 수 있도록이 커널을 어떻게 변경합니까?

cudamatrix-multiplication

답변 # 1

Matlab을 사용하고 있고 배열은 열 순서대로 저장되므로 커널 코드를 약간 수정해야합니다.

#include <thrust/iterator/counting_iterator.h>
#include <thrust/copy.h>
#include <thrust/device_vector.h>
#include <iostream>
__global__ void matrixMultiplication(const double *A, const double *B, double *C, int M, int N)
{
    int i = blockDim.y * blockIdx.y + threadIdx.y;
    int j = blockDim.x * blockIdx.x + threadIdx.x;
    double value = 0;
    for(int k = 0; k < N; k++){
        value += A[k * M + j] * B[i * M + k];
    }
    C[i * N + j] = value;
}
int main()
{
    const int M = 3, N = 4, K = 1;
    thrust::device_vector<double> A(M*N), B(N*K), C(M*K);
    thrust::counting_iterator<double> counter(1.0);
    thrust::copy(counter, counter + (M*N), A.begin());
    thrust::copy(counter, counter + (N*K), B.begin());
    dim3 grid(1,1), block(M,K);
    matrixMultiplication<<<grid, block>>>( thrust::raw_pointer_cast(A.data()),
                                           thrust::raw_pointer_cast(B.data()),
                                           thrust::raw_pointer_cast(C.data()),
                                           M, N );
    cudaDeviceSynchronize();
    for(int i=0; i<M*K; i++)
        std::cout << C[i] << std::endl;
    return 0;
}

커널에는 출력 매트릭스의 항목 당 하나의 스레드가 필요하므로 (3 x 4) * (4 x 1) 예제에서 3 x 1 스레드를 실행해야합니다. 실행하면 다음을 볼 수 있습니다 :

$ nvcc -arch=sm_52 -std=c++11 -o spoonfull spoonfull.cu 
$ ./spoonfull 
70
80
90

열 주요 주문 저장에 적합합니다.

Python2.net

nx1 벡터에서 mxn 행렬의 Cuda 곱셈

관련 자료