“CUDA”目录存档

cuda Array 拷贝问题

2010年05月3日,星期一

今天作者在使用cudaMemcpyArrayToArray函数时发现其效率极低,如果使用如下代码,需要77ms

cudaMemcpyArrayToArray(d_Array_A, 0, 0, d_Array_B, 0, 0, nFrameSize*sizeof(float));

如果开辟一块显存中转一下,效率会大幅提高,到20ms

cudaMemcpyFromArray(d_pfData, d_Array_B, 0,0,nFrameSize*sizeof(float),cudaMemcpyDeviceToDevice);
cudaMemcpyToArray(d_Array_A, 0, 0, d_pfData, nFrameSize*sizeof(float), cudaMemcpyDeviceToDevice);

纹理内存与全局内存之间有缓存,所以数据交换较快可以理解,但纹理内存之间拷贝速度反而更慢则不能理解。估计是cuda内部实现时没有做好优化。

Matlab下的CUDA编程(四)

2010年04月7日,星期三

第三节中我们介绍了NVIDIA的工程师写的编译脚本,但是这个脚本配置起来麻烦,用的时候选项也过于复杂。后来随着GPGPU的发展,CUDA逐渐被重视,因此MathWorks的工程师们重新写了这个nvmex脚本。毕竟是科班出身,新的脚本配置容易,使用简单。详细情况可访问其网站。在上述网站中,可找到下载nvmex源码的链接。下载后解压,得到nvmex.m。对其中的两个选项(红色部分按实际路径修改)稍加编辑,即可使用:

CUDA_LIB_Location = ‘C:\CUDA\lib’;
Host_Compiler_Location = ‘-ccbin "C:\Program Files\Microsoft Visual Studio 8\VC\bin"‘;

修改之后,将其复制到addMatrix.cu相同目录,并将此目录设置为matlab运行目录,在命令窗口输入:

>> nvmex(‘addMatrix.cu’);

即可完成编译。编译成功后,在matlab中即可像常规函数一样使用addMatrix函数。

Matlab下的CUDA编程(三)

2010年04月7日,星期三

根据第二节的介绍,matlab中可以通过mex文件的方式编译C/C++代码,但是对于.cu文件则无能为力。为了解决这个问题,NV的工程师们开发了用于编译cu文件的脚本,下面进行详细介绍。 测试环境:vs2005, matlab 7.6(r2008a), cuda 2.3 首先在NV的网站http://developer.nvidia.com/object/matlab_cuda.html下载matlab的插件包,并解压,共有4个重要文件:

nvmex.m
nvmex_helper.m
nvmexopts.bat
.\bin\nvmex.pl

首先将nvmex.pl拷贝到matlab的安装目录中的bin目录下,例如笔者的系统中就拷贝到C:\MATLAB\bin中。 下面编写cu文件,我们仍以两个矩阵相加为例,具体流程见代码注释。从代码中可看到,同时使用了matlab函数和cuda函数,并在GPU中完成了两个矩阵相加。完成addMatrix.cu的编写后,需要对其进行编译。 首先将nvmex.m,nvmex_helper.m,nvmexopts.bat拷贝到addMatrix.cu文件所在目录,并设置为matlab的当前目录。

  1. 在matlab命令窗口中运行: nvmex -setup,选择编译器,此处选择vs2005。
  2. 编辑nvmexopts.bat,修改其中的VSINSTALLDIR选项,指定为vs2005的安装目录,如笔者的系统中: C:\Program Files\Microsoft Visual Studio 8
  3. 编译命令: nvmex -f nvmexopts.bat addMatrix.cu -IC:\cuda\include -LC:\cuda\lib –lcudart

编译成功后,在matlab中即可像常规函数一样使用addMatrix函数。 可能会遇到的问题:

  1. 目录设置:需要仔细按照上述步骤设置好相关目录
  2. 编译时出现未定义的变量及结构,如_wchar等,则可更新nvmex.pl和nvmexopts.bat来解决。下载链接

(全文…)

CUDA 3.0正式发布

2010年03月20日,星期六

CUDA 3.0正式发布。下载链接。一些新的features(来自release notes,按照惯例稍微评论一下):

    新的API:

  • Half float(float16)纹理支持
    对一些精度要求不高的算法,可以copy更多的数据到显卡了,比如CT重见算法,已经有人写paper证明了:)
  • double3 double4 矢量类型支持
    没啥可说的了,就新的支持
  • 一维的设备到设备的拷贝支持streams了
    (确实没有感觉到stream的好处,之前折腾过一段,效果不明显)
  • 二进制ELF支持
  • Concurrent Kernels(通过CU_DEVICE_ATTRIBUTE_CONCURRENT_KERNELS 这个属性值才查看)
    貌似这个比较关键,也比较好,只是实际用起来不知道效果如何?
  • 批量的2D 3D FFT支持
    不做评论,暂时没有用到
    新的Toolkit功能增加

  • Nvcc的–host-compilation=C选项没有了
    这个没啥说的,难道这就是Nvidia说的原生的C++支持???
  • Window下cuda的DLL命名规定,也就是规范了下cudart.dll的名字,允许使用多个cudart了。比如cudart32_30_4.dll 就是32位3.0版的dll了,4是编译的个数??唉不会想那些vc运行库那样把
  • 将模拟器模式从cudart.dll单独分离出来了,以后用模拟器就要用cudaemu了
  • CUBLAS 新的函数
    没用到,不关注

(全文…)

Matlab下的CUDA编程(二)

2010年02月22日,星期一

Matlab下通过Mex文件编写C程序
本节参考NVIDIA网站相关资源,点击此处链接:

1、MEX规则

Matlab提供了MEX文件的方式来支持C/C++代码编写的算法。mex文件需要满足如下要求:
(1) 包含mex.h头文件
(2) 函数名称、参数返回值必须为如下形式:

void mexFunction(int nlhs, mxArray *plhs[],int nrhs, const mxArray *prhs[]);

其中:
nlhs 为输出数组个数(Left Hand Side)
plhs 为指向输出数组的指针
nrhs 为输入数组个数(Right Hand Side)
prhs 为指向输入数组的指针,且输入数组只读。
以上四个变量,是在C/C++代码中唯一可用到的变量。实际上,由于matlab中所有的变量都是mxArray结构(向量,数组,字符串。。。),因此常见的数据类型均可放入mxArray传递给程序进行处理。
(3)常用mex函数:
a. mex函数定义在mex.h(./extern/include)中,并以mex为前缀,例如打印输出的mexPrintf()函数、在mex文件中调用matlab函数的mexCallMATLAB()函数等。
b. 在matrix.h(./extern/include)中定义了mxArray结构以及对矩阵操作的函数,如创建双精度矩阵的mxCreateDoubleMatrix()、从mxArray中获取数据指针mxGetPr()等。
(4)编译
首先可以通过mex -setup来选择编译器,此时matlab会提示:
Would you like mex to locate installed compilers [y]/n?
此处如果选y,则会列出系统中安装的编译器,但不一定完整(作者就遇到这样的情况)。如果选n,则会列出matlab支持的所有编译器,
我们按实际情况选取即可(按照CUDA的要求,vs2005以上)。
另外常用的编译选项有:
-I 增加头文件(.h)包含目录
-L 增加库文件(.lib)包含目录
-l 链接引用的库文件名称

其他命令可查看matlab中的帮助文档。

2、MEX例程
在本教程中,举例实现将两个矩阵相加(C = A + B)。

/********************************************************************
filename:     addMatrix.c
file ext:       c
author:        wy@gpgpu.org.cn
purpose:  test matlab with c
*********************************************************************/
#include "mex.h"
void mexFunction(int nlhs, mxArray *plhs[],int nrhs, const mxArray *prhs[])
{
    int i, j, mA, nA, mB, nB, nMatSize;
    double *A, *B, *C;
    /* 输入变量检查 */
    if (nrhs != 2)
        mexErrMsgTxt("输入参数必须为2个");
    if (nlhs != 1)
        mexErrMsgTxt("输出参数必须为1个");
    mA = mxGetM(prhs[0]);
    nA = mxGetN(prhs[0]);
    mB = mxGetM(prhs[1]);
    nB = mxGetN(prhs[1]);
   if (mA != mB ||
       nA != nB)
       mexErrMsgTxt("输入矩阵尺寸必须相同");
    /*为输出结果分配内存*/
    plhs[0]=mxCreateDoubleMatrix(mA,nA,mxREAL);
    /*获取数据指针*/
    A = mxGetPr(prhs[0]);
    B = mxGetPr(prhs[1]);
    C = mxGetPr(plhs[0]);
    /*求和计算*/
    nMatSize = mA * nA;
    for (i=0; i<nmatsize ; i++)
    {
        C[i] = A[i] + B[i];
    }
    /*输出信息*/
    mexPrintf("求和完毕");
}

编译c文件:
mex -v addMatrix.c

编译成功后,在matlab中即可像常规函数一样使用addMatrix函数。
a = rand(2,2)
a =
0.93547      0.41027
0.9169      0.89365

b = rand(2,2)
b =
0.057891      0.81317
0.35287    0.0098613

c = addMatrix(a,b)
求和完毕
c =
0.99336       1.2234
1.2698      0.90351

3. 使用cuda标准库
在nvidia给出的白皮书例程中,给出了二维FFT的实现方法,本文将其引用,说明通过mex文件使用cuda标准库实现加速的方法。
在该程序中,没有调用内核程序,而只使用了runtime库和fft库,并不需要.cu文件,可以用mex命令直接编译。因此,将cuda算法封装之后再由mex文件调用是使用matlab使用cuda的一种方法。
编译命令:
mex -IC:\CUDA\include fft2_cuda.c -LC:\cuda\lib -lcufft -lcudart
(全文…)

Matlab下的CUDA编程(一)

2010年02月4日,星期四

引言

Matlab作为科学计算中的重要工具,它提供了丰富的基本函数和工具箱,从而在各个领域得到了广泛的应用。但Matlab的缺点是代码效率较低,在工程应用中只能作为模型验证,而不能得到实际应用。而CUDA作为显卡编程比较成熟的语言,能够充分利用GPU的计算能力,提高执行效率。因此如何能将CUDA的高效与matlab的简便有机的结合是本文要解决的问题,根据解决问题的方式我们将分为两节进行讲解。

一种简单的GPU三维图像分割算法

2009年12月6日,星期日

图像分割的定义是将整幅图像区域分割成各自互不交叉的小区域。图像分割的算法有很多,教科书上讲到一些经典的分割算法,包括阈值法、区域生长法、分裂与合并算法、边界跟踪与连接算法、分水岭算法,这些算法适合提取出灰度比较一致的各个区域;此外还有纹理分割算法,严格说来,与其称之为纹理分割还不如称之为纹理检测与分类,通过提取纹理特征,如:Laws特征、梯度共生矩阵特征、分形特征,得到每一个点或小邻域的特征向量,然后利用特征向量将每个点分到不同的纹理类中去。
有些扯远了,还是回到三维图像分割上来,由于三维图像的数据量是巨大的,暂不考虑复杂的纹理特征,我们仅仅希望有在GPU上实现一种并行的快速算法,能够自动标记出三维数据中所有灰度比较一致的各个连通区域。我们先看看教科书上的方法,阈值法不能提取多个灰度不一致的区域,区域生长不容易做到每个区域种子点自动选取,分裂与合并的逻辑稍显复杂且分割的区域边缘会出现锯齿,边界跟踪与连接在三维上根本没办法实现,分水岭又是串行递归的-_-!!!
假设已经是有了一幅二值图,我们就可以用连通区标记的算法,为二值图中的每一个小的区域标上不同的标号,这就是常用的二值图CCL(Connected Component Labeling)算法,可以扫描填充法、FloodFill等方式来具体实现这个算法。算法的原理是,判断当前点的邻域内有没有和当前点取值相同的点,如果有,就把他们标记为相同的标号。利用这个思想,如果在判断的时候,不是判断当前点的邻域内有没有和当前点取值相同的点,而是判断当前点的邻域内有没有和当前点取值相近的点——有点类似于梯度区域生长算法,就可以设计出适用于灰度图像的灰度连通区标记算法灰度图CCL算法,如果对每一个点进行如上判断,就可以预先生成一幅具有二值图,且对每个点的判断不依赖于其它点判断的结果,可以做到完全并行。这一段代码很简单,假设体数据已经做了去噪处理,然后用For循环遍历每一个点就能搞定(对于Cuda实现,就是把最里层的内容写到kernel里面),贴代码:
(全文…)

Ubuntu 9.04 安装 CUDA 开发环境

2009年11月13日,星期五

1. 下载 CUDA 文件

从 Nvidia CUDA 官方网站( http://www.nvidia.com/object/cuda_home.html )的下载页面可以找到适用于 ubuntu 9.04 的 CUDA 文件,包括:

*  NVIDIA Driver 190.18 Beta for Linux (Ubuntu 9.04) with CUDA Support
*  CUDA Toolkit 2.3 for Linux (Ubuntu 9.04)
*  CUDA SDK 2.3 code samples for Linux (Ubuntu 9.04) (全文…)

CUDA 3.0

2009年11月6日,星期五

The CUDA Toolkit 3.0 Beta is now available to GPU Computing registered
developers

CUDA上体数据的中值滤波算法讨论

2009年10月21日,星期三

(全文…)