矩阵乘法时遇到的问题

加速计算 CUDA
1

最近使用共享内存写了一个矩阵乘法的程序，kernel以及相关的host端程序如下，其中N代表矩阵的大小，blocksize*blockssize为每个block中的线程数

global void kernel(int a_d,int b_d,int c_d)
{
shared int a_share[BlockSizeBlockSize],b_share[BlockSizeBlockSize];
int t = 0;
int x = threadIdx.x+blockIdx.xblockDim.x;
int y = threadIdx.y+blockIdx.y*blockDim.y;

if (x+yN < NN)
{
a_share[threadIdx.x+threadIdx.yBlockSize] = a_d[x+yN];
b_share[threadIdx.x+threadIdx.yBlockSize] = b_d[x+yN];
}
else
{
a_share[threadIdx.x+threadIdx.yBlockSize] = 0;
b_share[threadIdx.x+threadIdx.yBlockSize] = 0;
}
__syncthreads();
for (int i = 0;i<BlockSize;i++)
{
t += a_share[i+threadIdx.y*BlockSize]b_share[threadIdx.x+iBlockSize];
}
c_d[x+y*N] = t;
}

host端相关定义如下
dim3 threadsPerBlock(BlockSize,BlockSize);
dim3 blockPerGrid((N-1+BlockSize)/BlockSize,(N-1+BlockSize)/BlockSize);
kernel<<<blockPerGrid,threadsPerBlock>>>(a_d,b_d,c_d);

现在问题在于，当N较小的时候，如2000以内，计算都没有问题，但当N大于3000时，计算会出错。而且由于已经关掉了TDR，有时候会出现假死的情况，即鼠标动不了，呼叫任务管理器也没有反应。请问斑竹这是怎么回事？

2

TDR 是什么，百度出一堆含义，感觉都不对。楼主科普一下？

3

LZ您好：

请提供您计算出错的具体情况，以便分析。

4

请看这里：

http://msdn.microsoft.com/en-us/library/windows/hardware/gg487368.aspx

5

这不很正常么。

当关闭了TDR后，您的kernel需要长时间运行，将不会自动终止。自然您遇到此时“假死”的状况是正常的。

也许贵卡上在N>3000的时候，kernel需要较长的运行时间，请不妨长期等待下（例如去睡个午觉），然后看看回来后是否正常完成。

感谢来访。

6

很好的科普源头，谢谢！

7

原来如此···那可以设置成使用集成显卡进行显示，然后N卡进行计算吗？这样是不是计算的同时就不会产生假死的状况？

8

楼主您好：

这个是可以的，但需要您的BIOS支持在插入独显的时候不会自动禁用集成显卡。
当上句满足的时候，您需要继续在BIOS里设置init display first为集显（或者类似字样），
然后您还需要在win7上将桌面设置到集成显卡上，
以及您还需要给集成显卡连接显示器。

这样您的独显就不用承担显示任务了。

:slight_smile:

9

但这不代表不会产生“假死”现象，如果例如你的kernel里有死循环，该死还是会死的。

您觉得呢？

10

[

额···我说的假死指的是直接鼠标都动不了···
我以为这是由于显卡忙于计算导致无法处理系统显示···所以才想说能不能使用集成显卡来承担显示任务。这样即便kernel里面是死循环，也不过导致显卡不断执行计算任务，而不致于连鼠标都动不了吧····

11

是的。

如果“假死”是您这么解释，那么无问题。
不过需要注意的是，这并不妨碍在您的另外一张独显上，您的kernel在死循环。
是吧。

12

恩，这个没问题，是我的措辞可能有些问题
多谢版主大大答疑解惑