‘壹’ 如何学习cuda c
1、CUDAC编写WindowsConsoleApplication
下面我们从一个简单的例子开始学习CUDAC。
打开VS,新建一个CUDAWinApp项目,项目名称为Vector,解决方案名称为CUDADemo。依次点击“确定”,“下一步”,选择Emptyproject。点击“Finished”。这样一个CUDA的项目就建成了。
右键点击Vector项目,依次选择“添加”、“新建项”、“代码”、“CUDA”。在名称中输入要添加的拿唯则文件名。如Vector.cu。然后点击添加。
下面在Vector.cu文件里实现两个向量相加的程序。
//添加系统库
#include
#include
//添加CUDA支持
#include
__global__voidVecAdd(float*A,float*B,float*C);
__host__voidrunVecAdd(intargc,char**argv);
intmain(intargc,char**argv)
{
runVecAdd(argc,argv);
CUT_EXIT(argc,argv);
}
__host__voidrunVecAdd(intargc,char**argv)
{//初始化host端内存数据
constunsignedintN=8;//向量维数
constunsignedintmemSize=sizeof(float)*N;//需要空间的字节数
float*h_A=(float*)malloc(memSize);
float*h_B=(float*)malloc(memSize);
float*h_C=(float*)malloc(memSize);
for(unsignedinti=0;i<N;i++)
{h_A[i]=i;h_B[i]山唯=i;}
//设备端显存空间
float*d_A,*d_B,*d_C;
//初始化Device
CUT_DEVICE_INIT(argc,argv);
CUDA_SAFE_CALL(cudaMalloc((void**)&d_A,memSize));
CUDA_SAFE_CALL(cudaMalloc((void**)&d_B,memSize));
CUDA_SAFE_CALL(cudaMalloc((void**)&d_C,memSize));
CUDA_SAFE_CALL(cudaMemcpy(d_A,h_A,memSize,cudaMemcpyHostToDevice));
CUDA_SAFE_CALL(cudaMemcpy(d_B,h_B,memSize,cudaMemcpyHostToDevice));
VecAdd<<<1,N,memSize>>>(d_A,d_B,d_C);
CUT_CHECK_ERROR("Kernelexecutionfailed");
CUDA_SAFE_CALL(cudaMemcpy(h_C,d_C,memSize,cudaMemcpyDeviceToHost));
for(unsignedinti=0;i<N;i++)
{printf("%.0f",h_C[i]);}
free(h_A);free(h_B);free(h_C);
CUDA_SAFE_CALL(cudaFree(d_A));
CUDA_SAFE_CALL(cudaFree(d_B));
CUDA_SAFE_CALL(cudaFree(d_C));
}
__global__voidVecAdd(float*A,float*B,float*C)
{
//分配sharedmemory
extern__shared__floats_A[];
extern__shared__floats_B[];
extern__shared__floats_C[];
//从globalmemory拷贝到sharedmemory
constunsignedinti=threadIdx.x;
s_A[i]=A[i];
s_B[i]=B[i];
//计算
s_C[i]=s_A[i]+s_B[i];
//拷贝到globalmemory
C[i]=s_C[i];
}
由于这里不是讲CUDA编程的,关于它的编程模型已经超出了我要介绍的范围,您可以阅读消棚《GPU高性能运算之CUDA》来获得CUDA编程模型的知识。
编译Vector项目,执行此项目后会得到图1如下输出:
图1Vector项目执行结果
2、CUDAC编写DLL模块
更多情况下的您的软件可能只是使用CUDA来实现一段程序的加速,这种情况下我们可以使用CUDAC编写DLL来提供接口。下面我们就将例1编译成DLL。
在刚才的CUDADemo解决方案目录下添加一个新的CUDA项目(当然您也可以重新建立一个解决方案)。项目名为VecAdd_dynamic。ApplicationType选为DLL,AdditionalOptions选择EmptyProject。
第一步,添加头文件,文件名最好与工程名同名,这样便于您的维护工作。这里我向项目中添加了VecAdd_dynamic.h,在此头文件中添加如下代码
#ifndef_VECADD_DYNAMIC_H_
#define_VECADD_DYNAMIC_H_
//并行计算N维向量的加法
__declspec(dllexport)voidVecAdd(float*h_A,float*h_B,float*h_C,intN);
#endif
第二步,添加cpp文件,文件名为VecAdd_dynamic.cpp,在此文件中添加如下代码
#include
#include"VecAdd_dynamic.h"
#ifdef_MANAGED
#pragmamanaged(push,off)
#endif
BOOLAPIENTRYDllMain(HMODULEhMole,DWORDul_reason_for_call,LPVOIDlpReserved)
{
returnTRUE;
}
#ifdef_MANAGED
#pragmamanaged(pop)
#endif
第三步,添加def文件,此文件的功能就是确保其它厂商的编译器能够调用此DLL里的函数。这一点非常关键,因为您的程序可能用到多个厂家的编译器。文件名为VecAdd_dynamic.def。向该文件中添加:
EXPORTS
VecAdd
第四步,添加cu文件,文件名为VecAdd_dynamic.cu。注意此文件最好直接添加到项目目录下,不要添加到源文件选项卡或其它已有的选项卡下。
在cu文件里添加如下代码,实现要导出的函数。
#include
#include
#include
#if__DEVICE_EMULATION__
boolInitCUDA(void)
{returntrue;}
#else
boolInitCUDA(void)
{
intcount=0;
inti=0;
cudaGetDeviceCount(&count);
if(count==0)
{
fprintf(stderr,"Thereisnodevice./n");
returnfalse;
}
for(i=0;i<count;i++)
{
cudaDevicePropprop;
if(cudaGetDeviceProperties(&prop,i)==cudaSuccess)
{
if(prop.major>=1)
{break;}
}
}
if(i==count)
{
fprintf(stderr,"ThereisnodevicesupportingCUDA./n");
returnfalse;
}
cudaSetDevice(i);
printf("CUDAinitialized./n");
returntrue;
}
#endif
__global__voidD_VecAdd(float*g_A,float*g_B,float*g_C,intN)
{
unsignedinti=threadIdx.x;
if(i<N)
{g_C[i]=g_A[i]+g_B[i];}
}
voidVecAdd(float*h_A,float*h_B,float*h_C,intN)
{
if(!InitCUDA())
{return;}
float*g_A,*g_B,*g_C;
unsignedintsize=N*sizeof(float);
CUDA_SAFE_CALL(cudaMalloc((void**)&g_A,size));
CUDA_SAFE_CALL(cudaMalloc((void**)&g_B,size));
CUDA_SAFE_CALL(cudaMalloc((void**)&g_C,size));
CUDA_SAFE_CALL(cudaMemcpy(g_A,h_A,size,cudaMemcpyHostToDevice));
CUDA_SAFE_CALL(cudaMemcpy(g_B,h_B,size,cudaMemcpyHostToDevice));
D_VecAdd<<<1,N>>>(g_A,g_B,g_C,N);
CUDA_SAFE_CALL(cudaMemcpy(h_C,g_C,size,cudaMemcpyDeviceToHost));
cudaFree(g_A);cudaFree(g_B);cudaFree(g_C);
}
第五步,如果您已经正确完成了以上四步,那么剩下的就只有编译,只要您用过VS,这一步就不需要我介绍了吧。成功之后,在您的解决方案文件目录下的Debug文件夹下会有一个VecAdd_dynamic.dll文件。
3、在.NET中使用CUDAC编写的DLL
下面介绍在托管程序中如何使用VecAdd_dynamic.dll。
第一步,在上面的解决方案CUDADemo下添加一个C++/CLR的Windows窗体应用程序,工程名为NETDemo(当然您也可以重新建一个解决方案,工程名也是随意的)。
第二步,在窗体上添加一个按钮,名字随意,我将它的现实文本改为“调用CUDA_DLL”,给这个按钮添加click事件。我们的代码将在这个事件里添加调用VecAdd()的程序。在窗体上添加一个文本框用来显示调用输出的结果。
第三步,代码实现。为工程NETDemo添加一个头文件,我将它命名为Win32.h,这个文件中主要是实现VecAdd()函数的导入。在此文件中添加如下代码
#pragmaonce
namespaceWin32
{
usingnamespaceSystem::Runtime::InteropServices;
[DllImport("VecAdd_dynamic.dll",EntryPoint="VecAdd",CharSet=CharSet::Auto)]
extern"C"voidVecAdd(float*h_A,float*h_B,float*h_C,intN);
}
在Form1.h中,#pragmaonce之后namespaceNETDemo之前添加以下代码。
#include"Win32.h"
#include
在button1_Click()中添加如下代码
intN=8;
float*h_A=(float*)malloc(N*sizeof(float));
float*h_B=(float*)malloc(N*sizeof(float));
float*h_C=(float*)malloc(N*sizeof(float));
for(inti=0;i<N;i++)
{h_A[i]=i;h_B[i]=i;}
Win32::VecAdd(h_A,h_B,h_C,N);
String^reslut;
for(inti=0;i<N;i++)
{reslut+=Convert::ToString(h_C[i])+",";}
this->textBox1->Text=Convert::ToString(reslut);
free(h_A);free(h_B);free(h_C);
第四步、执行NETDemo项目。点击“调用CUDA_DLL”,您会看到图3所示的结果
图3NETDemo运行结果
到现在为止您已经完全可以正确使用CUDA了。
‘贰’ 如何进行CUDA C程序核函数的调试
由于CUDA调试工具的不完善、CUDA调试工具上手难度较高,并行思想本身就难调试等因素,CUDA调试一直都是一件很蛋疼的事情。写CUDA也有三四年了,前段时间在群里见别人问CUDA调试的问题,突然有想法写个CUDA调试的博客。自己经验尚浅,希望各位大大看过后能够在评论里指点一二,共同完善这篇博客。
本博客只针对逻辑bug。
1 定位bug
出现bug的第一想法自然是定位bug。cuda比较奇特的地方在于,有时报错bug在500行,但500行出的代码没有错误,而是在1000行的地方逻辑错了,十分头疼。
下面介绍三种我总结的定位bug方法:
1.1 二分法
一半一半的注释代码,定位bug。比较笨拙和麻烦,但是十分好用。
1.2 输出定位法
将整体代码分为几个模块,正常的CUDA代码大概可以分为数据返纯初始化,内存申请,内存拷贝,核函数执行,结果拷贝等模块。在每个模块结束后输出标志,示例如图1。这样在调试时就可以根据输出快速定位bug大约在什么位置。如下图:
‘叁’ cuda并行程序设计 gpu编程指南 pdf
这本书非常好,绝对是CUDA编程的一个必不可少的工具书。
但是我建议楼主读原着的英文版的更好一点,中文版的翻译有的地方并不准确。
下面是下载文件:
望采纳
‘肆’ linux下CUDA程序一般怎么编译
有以下步骤:
1.源程序的编译
在Linux下面,如果要编译一个C语言源程序,我们要使用GNU的gcc编译器. 下面
我们以一个实例来说明如何使用gcc编译器.
假设我们有下面一个非常简单的源程序(hello.c):
int main(int argc,char **argv)
{
printf("Hello Linux\n");
}
要编译这个程序,我们只要在命令行下执行:
gcc -o hello hello.c
gcc 编译器就会为我们生成一个hello的可执行文件.执行./hello就可以看到程
序的输出结果了.命令行中 gcc表示我们是用gcc来编译我们的源程序,-o 选项表示
我们要求编译器给我们输出的可执行文件名为hello 而hello.c是我们的源程序文件.
gcc编译器有许多选项,一般来说我们只要知道其中的几个就够了. -o选项我们
已经知道了,表示我们要求输出的可执行文件名. -c选项表示我们只要求编译器输出
目标代码,而不必要输出可执行文件. -g选项表示我们要求编译器在编译的时候提
供我们以后对程序进行调试的信息.
知道了这三个选项,我们就可以编译我们自己所写的简单的源程序了,如果你
想要知道更多的选项,可以查看gcc的帮助文档,那里有着许多对其它选项的详细说
明.
2.Makefile的编写
假设我们有下面这样的一个程序,源代码如下:
#include "mytool1.h"
#include "mytool2.h"
int main(int argc,char **argv)
{
mytool1_print("hello");
mytool2_print("hello");
}
#ifndef _MYTOOL_1_H
#define _MYTOOL_1_H
void mytool1_print(char *print_str);
#endif
#include "mytool1.h"
void mytool1_print(char *print_str)
{
printf("This is mytool1 print %s\n",print_str);
}
#ifndef _MYTOOL_2_H
#define _MYTOOL_2_H
void mytool2_print(char *print_str);
#endif
#include "mytool2.h"
void mytool2_print(char *print_str)
{
printf("This is mytool2 print %s\n",print_str);
}
当然由于这个程序是很短的我们可以这样来编译
gcc -c main.c
gcc -c mytool1.c
gcc -c mytool2.c
gcc -o main main.o mytool1.o mytool2.o
这样的话我们也可以产生main程序,而且也不时很麻烦.但是如果我们考虑一
下如果有一天我们修改了其中的一个文件(比如说mytool1.c)那么我们难道还要重
新输入上面的命令?也许你会说,这个很容易解决啊,我写一个SHELL脚本,让她帮我
去完成不就可以了.是的对于这个程序来说,是可以起到作用的.但是当我们把事情
想的更复杂一点,如果我们的程序有几百个源程序的时候,难道也要编译器重新一
个一个的去编译?
为此,聪明的程序员们想出了一个很好的工具来做这件事情,这就是make.我们
只要执行以下make,就可以把上面的问题解决掉.在我们执行make之前,我们要先
编写一个非常重要的文件.--Makefile.对于上面的那个程序来说,可能的一个
Makefile的文件是:
# 这是上面那个程序的Makefile文件
main:main.o mytool1.o mytool2.o
gcc -o main main.o mytool1.o mytool2.o
main.o:main.c mytool1.h mytool2.h
gcc -c main.c
mytool1.o:mytool1.c mytool1.h
gcc -c mytool1.c
mytool2.o:mytool2.c mytool2.h
gcc -c mytool2.c
有了这个Makefile文件,不过我们什么时候修改了源程序当中的什么文件,我们
只要执行make命令,我们的编译器都只会去编译和我们修改的文件有关的文件,其
它的文件她连理都不想去理的.
下面我们学习Makefile是如何编写的.
在Makefile中也#开始的行都是注释行.Makefile中最重要的是描述文件的依赖
关系的说明.一般的格式是:
target: components
TAB rule
第一行表示的是依赖关系.第二行是规则.
比如说我们上面的那个Makefile文件的第二行
main:main.o mytool1.o mytool2.o
表示我们的目标(target)main的依赖对象(components)是main.o mytool1.o
mytool2.o 当倚赖的对象在目标修改后修改的话,就要去执行规则一行所指定的命
令.就象我们的上面那个Makefile第三行所说的一样要执行 gcc -o main main.o
mytool1.o mytool2.o 注意规则一行中的TAB表示那里是一个TAB键
Makefile有三个非常有用的变量.分别是$@,$^,$<代表的意义分别是:
$@--目标文件,$^--所有的依赖文件,$<--第一个依赖文件.
如果我们使用上面三个变量,那么我们可以简化我们的Makefile文件为:
# 这是简化后的Makefile
main:main.o mytool1.o mytool2.o
gcc -o $@ $^
main.o:main.c mytool1.h mytool2.h
gcc -c $<
mytool1.o:mytool1.c mytool1.h
gcc -c $<
mytool2.o:mytool2.c mytool2.h
gcc -c $<
经过简化后我们的Makefile是简单了一点,不过人们有时候还想简单一点.这里
我们学习一个Makefile的缺省规则
.c.o:
gcc -c $<
这个规则表示所有的 .o文件都是依赖与相应的.c文件的.例如mytool.o依赖于
mytool.c这样Makefile还可以变为:
# 这是再一次简化后的Makefile
main:main.o mytool1.o mytool2.o
gcc -o $@ $^
.c.o:
gcc -c $<
好了,我们的Makefile 也差不多了,如果想知道更多的关于Makefile规则可以查
看相应的文档.
3.程序库的链接
试着编译下面这个程序
#include
int main(int argc,char **argv)
{
double value;
printf("Value:%f\n",value);
}
这个程序相当简单,但是当我们用 gcc -o temp temp.c 编译时会出现下面所示
的错误.
/tmp/cc33Ky.o: In function `main':
/tmp/cc33Ky.o(.text+0xe): undefined reference to `log'
collect2: ld returned 1 exit status
出现这个错误是因为编译器找不到log的具体实现.虽然我们包括了正确的头
文件,但是我们在编译的时候还是要连接确定的库.在Linux下,为了使用数学函数,我
们必须和数学库连接,为此我们要加入 -lm 选项. gcc -o temp temp.c -lm这样才能够
正确的编译.也许有人要问,前面我们用printf函数的时候怎么没有连接库呢?是这样
的,对于一些常用的函数的实现,gcc编译器会自动去连接一些常用库,这样我们就没
有必要自己去指定了. 有时候我们在编译程序的时候还要指定库的路径,这个时候
我们要用到编译器的 -L选项指定路径.比如说我们有一个库在 /home/hoyt/mylib下
,这样我们编译的时候还要加上 -L/home/hoyt/mylib.对于一些标准库来说,我们没
有必要指出路径.只要它们在起缺省库的路径下就可以了.系统的缺省库的路径/lib
/usr/lib /usr/local/lib 在这三个路径下面的库,我们可以不指定路径.
还有一个问题,有时候我们使用了某个函数,但是我们不知道库的名字,这个时
候怎么办呢?很抱歉,对于这个问题我也不知道答案,我只有一个傻办法.首先,我到
标准库路径下面去找看看有没有和我用的函数相关的库,我就这样找到了线程
(thread)函数的库文件(libpthread.a). 当然,如果找不到,只有一个笨方法.比如我要找
sin这个函数所在的库. 就只好用 nm -o /lib/*.so|grep sin>~/sin 命令,然后看~/sin
文件,到那里面去找了. 在sin文件当中,我会找到这样的一行libm-2.1.2.so:00009fa0
W sin 这样我就知道了sin在 libm-2.1.2.so库里面,我用 -lm选项就可以了(去掉前面
的lib和后面的版本标志,就剩下m了所以是 -lm).
4.程序的调试
我们编写的程序不太可能一次性就会成功的,在我们的程序当中,会出现许许
多多我们想不到的错误,这个时候我们就要对我们的程序进行调试了.
最常用的调试软件是gdb.如果你想在图形界面下调试程序,那么你现在可以选
择xxgdb.记得要在编译的时候加入 -g选项.关于gdb的使用可以看gdb的帮助文件.由
于我没有用过这个软件,所以我也不能够说出如何使用. 不过我不喜欢用gdb.跟踪
一个程序是很烦的事情,我一般用在程序当中输出中间变量的值来调试程序的.当
然你可以选择自己的办法,没有必要去学别人的.现在有了许多IDE环境,里面已经自
己带了调试器了.你可以选择几个试一试找出自己喜欢的一个用.
5.头文件和系统求助
有时候我们只知道一个函数的大概形式,不记得确切的表达式,或者是不记得函数在那个头文件进行了说明.这个时候我们可以求助系统,比如说我们想知道fread这个函数的确切形式,我们只要执行 man fread 系统就会输出着函数的详细解释的.和这个函数所在的头文件说明了。如果我们要write这个函数说明,当我们执行man write时,输出的结果却不是我们所需要的。因为我们要的是write这个函数的说明,可是出来的却是write这个命令的说明。为了得到write的函数说明我们要用man 2 write。2表示我们用的是write这个函数是系统调用函数,还有一个我们常用的是3表示函数是c的库函数。
‘伍’ c语言 gpu
U越来越强大,GPU为显示图像做了优化之外,在计算上已经超越了通用的CPU。如此强大的芯片如果只是作为显卡就太浪费了,因此NVidia推出CUDA,让显卡可以用于图像计算以外的目的,也就是超于游戏,使得GPU能够发挥其强大的运算能力。
一年前NVIDIA发布CUDA,这是一种专门针对GPU的C语言开发工具。与以往采用图形API接口指挥GPU完成各种运算处理功能不同,CUDA的出现使研究人员和工程师可以在熟悉的C语言环境下,自由地输入代码调用GPU的并行处理架构。这使得原先需要花费数天数周才能出结果的运算大大缩短到数几小时,甚至几分钟之内。
CUDA是用于GPU计算的开发环境,它是一个全新的软硬件架构,可以将GPU视为一个并行数据计算的设备,对所进行的计算进行分配和管理。在CUDA的架构中,这些计算不再像过去所谓的GPGPU架构那样必须将计算映射到图形API(OpenGL和Direct 3D)中,因此对于开发者来说,CUDA的开发门槛大大降低了。CUDA的GPU编程语言基于标准的C语言,因此任何有C语言基础的用户都很容易地开发CUDA的应用程序。
那么,如何使得CPU与GPU之间很好的进行程序之间的衔接呢?以GPGPU的概念来看,显卡仍然需要以传统的DirectX和OpenGL这样的API来实现,对于编程人员来说,这样的方法非常繁琐,而CUDA正是以GPGPU这个概念衍生而来的新的应用程序接口,不过CUDA则提供了一个更加简便的方案——C语言。我们回顾一下CUDA的发展历史。
‘陆’ CUDA编程基础——Grid、Block、Thread
本消尺余文主要介绍三者之间的关系。
三者之间关系如图所示,从中可以看出,三者存在包含关系。每个grid分为多个block,每困枣个block分为多个Thread,grid和block最多可拿滚以是三维的。
‘柒’ NVIDIA显卡支持CUDA,什么是CUDA
关于CUDA:
CUDA(Compute Unified Device Architecture)是一个新的基础架构,这个架构可以使用GPU来解决商业、工业以及科学方面的复杂计算问题。它是一个完整的GPGPU解决方案,提供了硬件的直接访问接口,而不必像传统方式一样必须依赖图形API接口来实现GPU的访问。在架构上采用了一种全新的计算体系结构来使用GPU提供的硬件资源,从而给大规模的数据计算应用提供了一种比CPU更加强大的计算能力。CUDA采用C语言作为编程语言提供大量的高性能计算指令开发能力,使开发者能够在GPU的强大计算能力的基础上建立起一种效率更高的密集数据计算解决方案。
关于NVIDIA CUDA技术
NVIDIA CUDA技术是当今世界上唯一针对NVIDIA GPU(图形处理器)的C语言环境,为支持CUDA技术的NVIDIA GPU(图形处理器)带来无穷的图形计算处理性能。凭借NVIDIA CUDA技术,开发人员能够利用NVIDIA GPU(图形处理器)攻克极其复杂的密集型计算难题,应用到诸如石油与天然气的开发,金融风险管理,产品设计,媒体图像以及科学研究等领域。
CUDA™ 工具包是一种针对支持CUDA功能的GPU(图形处理器)的C语言开发环境。CUDA开发环境包括:
nvcc C语言编译器
适用于GPU(图形处理器)的CUDA FFT和BLAS库
分析器
适用于GPU(图形处理器)的gdb调试器(在2008年3月推出alpha版)
CUDA运行时(CUDA runtime)驱动程序(目前在标准的NVIDIA GPU驱动中也提供)
CUDA编程手册
CUDA开发者软件开发包(SDK)提供了一些范例(附有源代码),以帮助使用者开始CUDA编程。这些范例包括:
并行双调排序
矩阵乘法
矩阵转置
利用计时器进行性能评价
并行大数组的前缀和(扫描)
图像卷积
使用Haar小波的一维DWT
OpenGL和Direct3D图形互操作示例
CUDA BLAS和FFT库的使用示例
CPU-GPU C—和C++—代码集成
二项式期权定价模型
Black-Scholes期权定价模型
Monte-Carlo期权定价模型
并行Mersenne Twister(随机数生成)
并行直方图
图像去噪
Sobel边缘检测滤波器
MathWorks MATLAB® 插件 (点击这里下载)
新的基于1.1版CUDA的SDK 范例现在也已经发布了。要查看完整的列表、下载代码,请点击此处。
技术功能
在GPU(图形处理器)上提供标准C编程语言
为在支持CUDA的NVIDIA GPU(图形处理器)上进行并行计算而提供了统一的软硬件解决方案
CUDA兼容的GPU(图形处理器)包括很多:从低功耗的笔记本上用的GPU到高性能的,多GPU的系统。
支持CUDA的GPU(图形处理器)支持并行数据缓存和线程执行管理器
标准FFT(快速傅立叶变换)和BLAS(基本线性代数子程序)数值程序库
针对计算的专用CUDA驱动
经过优化的,从中央处理器(CPU)到支持CUDA的GPU(图形处理器)的直接上传、下载通道
CUDA驱动可与OpenGL和DirectX图形驱动程序实现互操作
支持Linux 32位/64位以及Windows XP 32位/64位 操作系统
为了研究以及开发语言的目的,CUDA提供对驱动程序的直接访问,以及汇编语言级的访问。