关于矩阵相乘

大家好:
我刚刚开始学习cuda,希望从最基础的矩阵相乘开始。
我写了一个在CPU上的矩阵相乘的C语言程序(附件),MN=P,M是一个48的矩阵,N是一个8*2的矩阵。以二维数组的形式是可以实现的。
但我看关于CUDA的书籍,是将矩阵以一维的形式存储的(我是这么理解的,不知道对不对),为了用threadIdx.x和 threadIdx.y,以及M,N的尺寸进行矩阵元素的寻址。
在解决实际问题时,我还需要将二维的矩阵转化为一维的矩阵吗?
如果不用转化为一维矩阵,我需要如何实现矩阵元素的寻址。(我的思路是用指针实现,但每次还要将矩阵给个指针。有没有其它的方法。)
谢谢。

# include "stdio.h"

int main(void)
{        
   // part 1
   // allocate and initialize the matrices M,N,P
   // I/O to rea dthe input matries M and N
   printf("\n--------- this is part 1------------\n");
   // 4 matrix 
   int M_row=4;
   int M_col=8;
   int N_row=M_col;
   int N_col=2;
   
   // 4 for ite
   int i=0;
   int j=0;
   int k=0;
   float tmp=0;

   float M[M_row][M_col];
   float N[N_row][N_col]; 
   float P[M_row][N_col];

   // init M 
   printf("M=\n");
   for(i=0;i<M_row;i++)
   {
   for (j=0;j<M_col;j++)
   {                       
   M[i][j]=(i+1)+0.01*(j+1);
   printf("%3.3f,   ",M[i][j]);
   }
   printf("\n");
   }
   
   // init N 
   printf("N=\n");
   for(i=0;i<N_row;i++)
   {
   for (j=0;j<N_col;j++)
   {                       
   N[i][j]=(i+1)+0.001*(j+1);
   printf("%3.3f,   ",N[i][j]);
   }
   printf("\n");
   }
   
   // part 2
   // Matrix multiplic on CPU
   printf("\n ------------this is part 2------------\n");
   // cal P
   printf("M * N = P\n");
   printf("\n");
   for (i=0;i<M_row;i++)
   {
   for (j=0;j<N_col;j++)
   {
   tmp=0;
   for (k=0;k<M_col;k++)
   {
   tmp=tmp+M[i][k]*N[k][j];
   }
   P[i][j]=tmp;
   }
   }
   printf("\n--------------this is part 3 -----------\n");
   for (i=0;i<M_row;i++)
   {
   for (j=0;j<N_col;j++)
   {
   printf("%3.3f, ",P[i][j]);
   }
   printf("\n");
   }
   return 0;
}
//mod : Tue Apr  9 14:48:05 HKT 2013
//mod : Tue Apr  9 07:06:19 CST 2013
//mod :Mon Apr  8 06:53:31 CST 2013

楼主您好,您的问题问的好。

的确,我们可以使用float a[M][N];的形式,那么为何几乎所有的例子中都是用一维数组,然后进行坐标变换呢?

那是因为,float a[M][N]这种,在CUAD C/C++和它配套的host compiler VC中,均需要要求M, N是编译时刻确定的常数。

而一旦M,N是变量,则您无法使用。(需要C99标准的VLA支持,但cuda c和vc均不支持VLA)。

所以,除了固定的矩阵大小的情况,任何为了能处理通用的矩阵的乘法,您只能写成一维的,然后将宽度和高度信息传入,然后自己计算。(请注意,这并不会带来额外的性能损失)。

所以,您的疑惑就此解开。
以及,不排除以后cuda c支持VLA(以及其指针形式变种), 以后可能您可以直接使用,例如:
global void meatball(float *p, int width, int height)
{
float (array)[width] = (float ()[width])p; //建立指针,width是运行时才可知的
//以下直接使用:
//array[ j ] [i ]形式
}

该操作需要未来版本的cuda toolkit支持。您目前还不能这么用。(只是给您提前说明)。

感谢“横扫千军”的回复。
学习了。可能还需要再仔细想想(因为比较菜)。但在大体的意思已经明白了。

您客气了,服务您是我的荣幸。