今天本来想写一个矩阵的卷积运算的,但是果真自己动手写代码就遇到各种问题了。
本来想在GPU中生成一个下面的Matrix结构体,再让GPU做运算。
structMatrix{float* dptr;introw;intcol;}
初始代码:
Matrix*m;HANDLE_ERROR(cudaMalloc( (void**)&m,sizeof(Matrix)) );
想了一下,这样应该在GPU中存在Matrix结构体了,但是里面的指针dptr是野指针阿。
那怎么办?
要么在cpu代码中为m->dptr在gpu申请内存并返回地址,
要么直接在gpu代码中为m->dptr在gpu申请内存并返回地址。
第一个思路:
在cpu代码中为m->dptr在gpu申请内存并返回地址。
看起来不靠谱,因为此时的m的结构体在gpu内存中的,在cpu代码无从触及,就是无法进行m->dptr的操作。
得绕一下,通过拷贝的形式。
参考了这个链接的答案:
http://stackoverflow.com/questions/19404965/how-to-use-cudamalloc-cudamemcpy-for-a-pointer-to-a-structure-containing-point
具体看链接,这里简单提一下。
其实就是现在cpu内存中生成一个m,然后把m->dptr指向GPU申请的内存。再把这个m从cpu内存中拷贝到gpu中。
当时我机器拷贝又遇到了问题,拷贝结构体失败了。当初怀疑是不是不能给gpu传结构体。拿了http://stackoverflow.com/questions/5984460/copying-struct-data-from-host-to-device-on-cuda-using-cudamemcpy的代码来跑一下,发现出错。
后来改了一下编译参数,发现好了。
把-gencodearch=compute_20,code=sm_20改为
-gencodearch=compute_12,code=sm_12。
对编译参数的理解。
-gencodearch=compute_20,code=sm_20
这个参数告诉编译器,以哪个计算能力版本的标准来编译代码,比如上面这个就是20。
如果机器的实际版本是12,而用上面的参数让编译器去编译,则可能编译没有错误,但运行有错。
所以,最好的做法就是,选择跟GPU计算能力就用对应的参数去编译。
第二个思路:
直接在gpu代码中为m->dptr在gpu申请内存并返回地址。
开始的疑问是,之前只见过在cpu代码中申请内存,没见过在gpu代码中申请内存,那能否这样呢?
参考了这个链接http://stackoverflow.com/questions/13480213/how-to-dynamically-allocate-arrays-inside-a-kernel
其中有这么一句话
Dynamicmemory allocation is only supported on compute capability 2.x and 3.xhardware.
意思就是在计算能力为2.x以上的GPU才支持在GPU代码中动态申请内存。
我的机器是1.2的,所以就不支持了。
但是还天真的试了一下用
-gencodearch=compute_20,code=sm_20编译参数,
发现编译没错,但是执行错了。所以,记得编译参数别乱写哦。
所以,现在的机器就只能用第一个思路了。
http://stackoverflow.com/questions/13480213/how-to-dynamically-allocate-arrays-inside-a-kernel
链接提到,在gpu代码中动态申请内存会影响性能。
后来又想了一下,如果直接在GPU代码中可以动态申请内存而且又不会影响性能,直接在GPU代码中newMatrix不就好了么?