1

35ページの『CUDA Cプログラミングガイド』の行列乗算の例を実行しています。練習のために、コードをコピーして、不足しているコードを完成させました。プログラムのロジックとその動作方法は理解していますが、期待どおりの結果が得られません。

これが私が作成した完全なコードです、エラーが私のものなのか、例によるものなのかわかりませんか？

コード： 

#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include "cuda_runtime.h"
#include "device_launch_parameters.h"
#include <iostream>    
#include <stdio.h>
#include <stdio.h>

using namespace std;
#define BLOCK_SIZE 16

typedef struct
{
    int width;
    int height;
    float *elements;
}Matrix;

__global__ void MatMulKernel(const Matrix,const Matrix, Matrix C);

void MatMul(const Matrix A,const Matrix B, Matrix C) 
{
    size_t size;
    //Matrix A creation y storage in device memory 
    Matrix d_A;
    d_A.width=A.width;
    d_A.height=A.height;
    size=A.height*A.width*sizeof(float);
    cudaMalloc(&d_A.elements,size);
    cudaMemcpy(d_A.elements,A.elements,size,cudaMemcpyHostToDevice);
    //Matrix B creation y storage in device memory 
    Matrix d_B;
    d_B.width=B.width;
    d_B.height=B.height;
    size=B.height*B.width*sizeof(float);
    cudaMalloc(&d_B.elements,size);
    cudaMemcpy(d_B.elements,B.elements,size,cudaMemcpyHostToDevice);
    //Matrix C creation y storage in device memory         
    Matrix d_C;
    d_C.width=C.width;
    d_C.height=C.height;
    size=C.height*C.width*sizeof(float);
    cudaMalloc(&d_C.elements,size);
    //        
    dim3 dimBlock(BLOCK_SIZE,BLOCK_SIZE);
    dim3 dimGrid(B.width/dimBlock.x,A.height/dimBlock.y);
    MatMulKernel<<<dimGrid,dimBlock>>>(d_A,d_B,d_C);
    //Copy the result in the matrix C from the device to the host.        
    cudaMemcpy(C.elements,d_C.elements,size,cudaMemcpyDeviceToHost);  
    //edit the missing code.
    // for(int i=0;i<BLOCK_SIZE*BLOCK_SIZE;i++){cout<<C.elements[i]<<endl;}      
    // result in random numbers
    cudaFree(d_A.elements);
    cudaFree(d_B.elements);
    cudaFree(d_C.elements);
}

__global__ void MatMulKernel(Matrix A, Matrix B, Matrix C)
{
    float Cvalue=0;
    int row=blockIdx.y*blockDim.y+threadIdx.y;
    int col=blockIdx.x*blockDim.x+threadIdx.x;
    for(int e=0;e<A.width;++e)
    {
        Cvalue+=A.elements[row*A.width+e]*B.elements[e*B.width+col];
    }
    C.elements[row*C.width+col]=Cvalue;
}

int main()
{
    cout<<"Matrices"<<endl;
    //Declarationd of the A,B,C matrix´s
    float a[15][15];        
    float b[15][15];
    float c[15][15];
    //Fill the matrix whit some numbers.
    int cont0=0;
    for(int c=0;c<15;c++)
    {
        for(int v=0;v<15;v++)
        {
            a[v][c]=cont0;
            b[v][c]=cont0;
            cont0++;
        }
    }
    //Flatten the matrix for the passing to the kernel
    int offset=0;
    float a_t[256];
    float b_t[256];
    for(int y=0;y<15;y++)
    {                        
        for(int x=0;x<15;x++)
        {
            a_t[x+offset]=a[x][y];
            b_t[x+offset]=a[x][y];
        }
        offset=offset+15;
    }
    float t_C[256];
    //Completing the matrix format for the kernel.
    Matrix m_A;
    m_A.height=15;
    m_A.width=15;
    m_A.elements=a_t;
    Matrix m_B;
    m_B.height=15;
    m_B.width=15;
    m_B.elements=b_t;
    Matrix m_C;
    m_C.height=15;
    m_C.width=15;
    m_C.elements=t_C;
    //Passing the formated matrix to the kernel.
    MatMul(m_A,m_B,m_C);                
    cout<<"Final"<<endl;        
return 0;
}

C.elementsプログラムはコンパイルされて実行されますが、 ：から の結果行列cudaMemcpy(C.elements,d_C.elements,size,cudaMemcpyDeviceToHost); は乱数です。配列へのポインタのように使用しようとしましたが、何も取得できず、配列のように扱うこともできません。

誰かが私がこれを終えるのを手伝ってくれるなら、私はうれしいです。

4

2 に答える 2

3

あなたのコードは、カーネルでの配列のインデックス付けと CPU での初期化の間にマイナーなミスマッチがあります。@harrism によって提案されたデバッグを含む修正されたコードは次のとおりです。

    #include "cuda_runtime.h"
    #include "device_launch_parameters.h"
    #include "cuda_runtime.h"
    #include "device_launch_parameters.h"
    #include <iostream>
    #include <stdio.h>
    #include <stdio.h>

    using namespace std;
    #define BLOCK_SIZE 16

    typedef struct
    {
        int width;
        int height;
        float *elements;
    }Matrix;

    __global__ void MatMulKernel(const Matrix,const Matrix, Matrix C);

    void MatMul(const Matrix A,const Matrix B, Matrix C)
    {
        size_t size;
        //Matrix A creation y storage in device memory
        Matrix d_A;
        d_A.width=A.width;
        d_A.height=A.height;
        size=A.height*A.width*sizeof(float);
        cudaMalloc(&d_A.elements,size);
        cudaMemcpy(d_A.elements,A.elements,size,cudaMemcpyHostToDevice);
        //Matrix B creation y storage in device memory
        Matrix d_B;
        d_B.width=B.width;
        d_B.height=B.height;
        size=B.height*B.width*sizeof(float);
        cudaMalloc(&d_B.elements,size);
    cudaMemcpy(d_B.elements,B.elements,size,cudaMemcpyHostToDevice);
    //Matrix C creation y storage in device memory
    Matrix d_C;
    d_C.width=C.width;
    d_C.height=C.height;
    //cudaMalloc(&d_C,sizeof(Matrix));
    //cudaMemcpy(d_C,C,sizeof(Matrix),cudaMemcpyHostToDevice);
    size=C.height*C.width*sizeof(float);
    cudaMalloc(&d_C.elements,size);
    //
    dim3 dimBlock(BLOCK_SIZE,BLOCK_SIZE);
    dim3 dimGrid(B.width/dimBlock.x,A.height/dimBlock.y);
    MatMulKernel<<<dimGrid,dimBlock>>>(d_A,d_B,d_C);
    //Copy the result in the matrix C from the device to the host.
    printf("error code: %s\n",cudaGetErrorString(cudaGetLastError()));
    cudaMemcpy(C.elements,d_C.elements,size,cudaMemcpyDeviceToHost);
    //
    cudaFree(d_A.elements);
    cudaFree(d_B.elements);
    cudaFree(d_C.elements);
}

__global__ void MatMulKernel(Matrix A, Matrix B, Matrix C)
{
        //printf("%d\n",threadIdx.x);
    float Cvalue=0;
    int row=blockIdx.y*blockDim.y+threadIdx.y;
    int col=blockIdx.x*blockDim.x+threadIdx.x;
    for(int e=0;e<A.width;++e)
    {
        Cvalue+=A.elements[row*A.width+e]*B.elements[e*B.width+col];
    }
    C.elements[row*C.width+col]=Cvalue;
}

int print_matrix(Matrix A){
        printf("Matrix:\n");
        int i;
        for(i=0; i<A.width*A.height; i++){
                if(i%A.width==0) printf("\n");
                printf("%6.4f\t",A.elements[i]);
        }
        printf("\n");
}
int main()
{
    cout<<"Matrices"<<endl;
    //Declarationd of the A,B,C matrix.s
    float a[BLOCK_SIZE][BLOCK_SIZE];
    float b[BLOCK_SIZE][BLOCK_SIZE];
    float c[BLOCK_SIZE][BLOCK_SIZE];
    //Fill the matrix whit some numbers.
    int cont0=0;
    for(int c=0;c<BLOCK_SIZE;c++)
    {
        for(int v=0;v<BLOCK_SIZE;v++)
        {
            a[v][c]=cont0;
            b[v][c]=cont0;
            cont0++;
        }
    }
    //Flatten the matrix for the passing to the kernel
    int offset=0;
    float a_t[BLOCK_SIZE*BLOCK_SIZE];
    float b_t[BLOCK_SIZE*BLOCK_SIZE];
    for(int y=0;y<BLOCK_SIZE;y++)
    {
        for(int x=0;x<BLOCK_SIZE;x++)
        {
            a_t[x+offset]=a[x][y];
            b_t[x+offset]=a[x][y];
        }
        offset=offset+BLOCK_SIZE;
    }
    float t_C[BLOCK_SIZE*BLOCK_SIZE];
    //Completing the matrix format for the kernel.
    Matrix m_A;
    m_A.height=BLOCK_SIZE;
    m_A.width=BLOCK_SIZE;
    m_A.elements=a_t;
    Matrix m_B;
    m_B.height=BLOCK_SIZE;
    m_B.width=BLOCK_SIZE;
    m_B.elements=b_t;
    Matrix m_C;
    m_C.height=BLOCK_SIZE;
    m_C.width=BLOCK_SIZE;
    m_C.elements=t_C;
    //Passing the formated matrix to the kernel.
    print_matrix(m_A);
    print_matrix(m_B);
    MatMul(m_A,m_B,m_C);
    print_matrix(m_C);
    cout<<"Final"<<endl;
return 0;
}

出力を確認します。結果が間違っている場合は、出力で報告されているシステムのカーネル エラーを確認してください。

于 2012-09-13T08:19:46.123 に答える
1

まず、質問に対する有用な回答を得る方法については、こちらを参照してください。特に、CUDA API 呼び出しとカーネル起動の戻り値を常に確認する必要があります。また、実行cuda-memcheckは、このような境界外アクセスを検出するのに非常に役立つことがよくあります。

@harrism は、結果に対して何もしていないように見えるのに、結果が間違っていることをどのように知っているかを尋ねました。

しかし、もっと重要なのは、16x16 のスレッドブロックで計算されている 15x15 の行列があるのに、範囲外のスレッドを無効にするように注意していないことです。簡単な例を作成しようとしているので、行列のサイズを 16x16 に増やします。奇数のサイズを処理する場合は、制御ロジックを実装する (または cuBLAS を使用する!) 必要があります。

于 2012-09-13T08:07:37.423 に答える