租房网站建设,网站怎么黑,wordpress 转发标题,登录qq网页版根据网站资料#xff0c;简单地汇编一下CUDA与OpenCL的区别。如有错误请指出。
题外话#xff1a; 美国Sandia国家实验室一项模拟测试证明#xff1a;由于存储机制和内存带宽的限制#xff0c;16核、32核甚至64核处理器对于超级计算机来说#xff0c;不仅不能带来性能提升…根据网站资料简单地汇编一下CUDA与OpenCL的区别。如有错误请指出。
题外话 美国Sandia国家实验室一项模拟测试证明由于存储机制和内存带宽的限制16核、32核甚至64核处理器对于超级计算机来说不仅不能带来性能提升甚至可能导致效率的大幅度下降。
什么是OpenCL
是由苹果Apple公司发起业界众多著名厂商共同制作的面向异构系统通用目的并行编程的开放式、免费标准也是一个统一的编程环境。便于软件开发人员为高性能计算服务器、桌面计算系统、手持设备编写高效轻便的代码而且广泛适用于多核心处理器(CPU)、图形处理器(GPU)、Cell类型架构以及数字信号处理器(DSP)等其他并行处理器在游戏、娱乐、科研、医疗等各种领域都有广阔的发展前景。
什么是CUDA
CUDA是一个基于Nvidia GPU的并行计算的架构。CUDA最主要的包含两个方面一个是ISA指令集架构第二硬件计算引擎实际上是硬件和指令集。 也就是说我们可以把CUDA看做是与X86或者cell类似的架构但是是基于是GPU而不是传统的CPU。 OpenCL实际上是什么
OpenCL实际上是针对异构系统进行并行编程的一个全新的APIOpenCL可以利用GPU进行一些并行计算的工作。 OpenGL是针对图形的而OpenCL则是针对并行计算的API。 OpenCL开发的过程中技术平台均为NVIDIA的GPU实际上OpenCL是基于NVIDIA GPU的平台进行开发的。另外OpenCL的第一次演示也是运行在NVIDIA的GPU上。从本质上来说OpenCL就是一个相当于Windows平台中DirectX那样的技术。或者说它是一个连接硬件和软件的API接口。在这一点上它和OpenGL类似不过OpenCL的涉及范围要比OpenGL大得多它不仅是用来作用于3D图形。如果用一句话描述OpenCL的作用就是通过调用处理器和GPU的计算资源释放硬件潜力让程序运行得更快更好。
CUDA实际上是什么
CUDA架构是原生的专门为计算接口而建造的这样的一个架构这种硬件架构包括指令集都是非常适合于这种并行计算为异构计算而设计的一整套的架构。CUDA架构可以支持API包括OpenCL或者DirectX同时CUDA还支持C、C语言还包括Fortran、Java、Python等各种各样的语言。 OpenCL与CUDA的关系是什么
CUDA和OpenCL的关系并不是冲突关系而是包容关系。OpenCL是一个API在第一个级别CUDA架构是更高一个级别在这个架构上不管是OpenCL还是DX11这样的API还是像C语言、Fortran、DX11计算都可以支持。作为程序开发员来讲一般他们只懂这些语言或者API可以采用多种语言开发自己的程序不管他选择什么语言只要是希望调用GPU的计算能在这个架构上都可以用CUDA来编程。关于OpenCL与CUDA之间的技术区别主要体现在实现方法上。基于C语言的CUDA被包装成一种容易编写的代码因此即使是不熟悉芯片构造的科研人员也可能利用CUDA工具编写出实用的程序。而OpenCL虽然句法上与CUDA接近但是它更加强调底层操作因此难度较高但正因为如此OpenCL才能跨平台运行。CUDA是一个并行计算的架构包含有一个指令集架构和相应的硬件引擎。OpenCL是一个并行计算的应用程序编程接口API在NVIDIA CUDA架构上OpenCL是除了C for CUDA外新增的一个CUDA程序开发途径。如果你想获得更多的对硬件上的控制权的话你可以使用OpenCL这个API来进行编程如果对API不是太了解也可以用CUDA C语言来编程这是两种不同编程的方式他们有他们相同点和不同点。但是有一点OpenCL和CUDA C语言进行开发的时候在并行计算这块他们的概念是差不多的这两种程序在程序上是有很大的相似度所以程序之间的相互移植相对来说也是比较容易。 CUDA C语言与OpenCL的定位不同或者说是用人群不同。CUDA C是一种高级语言那些对硬件了解不多的非专业人士也能轻松上手而OpenCL则是针对硬件的应用程序开发接口它能给程序员更多对硬件的控制权相应的上手及开发会比较难一些。程序员的使用习惯也是非常重要的一方面那些在X86 CPU平台使用C语言的人员会很容易接受基于CUDA GPU平台的C语言而习惯于使用OpenGL图形开发的人员看到OpenCL会更加亲切一些在其基础上开发与图形、视频有关的计算程序会非常容易。 给点实例看看
OpenCL 教學 http://www.kimicat.com/opencl-1/opencl-jiao-xue-yi 在 Windows 下使用 OpenCL http://www.kimicat.com/opencl-1/zai-windows-xia-shi-yong-opencl 在 Xcode 中使用 OpenCL http://www.kimicat.com/opencl-1/zai-xcode-zhong-shi-yong-opencl CUDA 教学 http://www2.kimicat.com/cuda%E7%B0%A1%E4%BB%8B 从很多方面来看CUDA和OpenCL的关系都和DirectX与OpenGL的关系很相像。如同DirectX和OpenGL一样CUDA和OpenCL中前者是配备完整工具包、针对单一供应商(NVIDIA)的成熟的开发平台后者是一个开放的标准。 虽然两者抱着相同的目标通用并行计算。但是CUDA仅仅能够在NVIDIA的GPU硬件上运行而OpenCL的目标是面向任何一种Massively Parallel Processor期望能够对不同种类的硬件给出一个相同的编程模型。由于这一根本区别二者在很多方面都存在不同 1开发者友好程度。CUDA在这方面显然受更多开发者青睐。原因在于其统一的开发套件(CUDA Toolkit, NVIDIA GPU Computing SDK以及NSight等等)、非常丰富的库(cuFFT, cuBLAS, cuSPARSE, cuRAND, NPP, Thrust)以及NVCC(NVIDIA的CUDA编译器)所具备的PTX(一种SSA中间表示为不同的NVIDIA GPU设备提供一套统一的静态ISA)代码生成、离线编译等更成熟的编译器特性。相比之下使用OpenCL进行开发只有AMD对OpenCL的驱动相对成熟。 2跨平台性和通用性。这一点上OpenCL占有很大优势这也是很多National Laboratory使用OpenCL进行科学计算的最主要原因。OpenCL支持包括ATI,NVIDIA,Intel,ARM在内的多类处理器并能支持运行在CPU的并行代码同时还独有Task-Parallel Execution Mode能够更好的支持Heterogeneous Computing。这一点是仅仅支持数据级并行并仅能在NVIDIA众核处理器上运行的CUDA无法做到的。 3市场占有率。作为一个开放标准缺少背后公司的推动OpenCL显然没有占据通用并行计算的主流市场。NVIDIA则凭借CUDA在科学计算、生物、金融等领域的推广牢牢把握着主流市场。再次想到OpenGL和DirectX的对比不难发现公司推广的高效和非盈利机构/标准委员会的低效(抑或谨慎想想C0x)。 我接触的很多开发者(包括我本人)都认为由于目前独立显卡市场的萎缩、新一代处理器架构(AMD的Graphics Core Next (GCN)、Intel的Sandy Bridge以及Ivy Bridge)以及新的SIMD编程模型(Intel的ISPC等)的出现未来的通用并行计算市场会有很多不确定因素CUDA和OpenCL都不是终点我期待未来会有更好的并行编程模型的出现(当然也包括CUDA和OpenCL如果它们能够持续发展下去)。 1. CUDA有远好于OpenCL的生态系统更易用对程序员更友好。OpenCL的API设计怪异缺乏一致性功能亦不正交很不直观远未成熟。 2. OpenCL的portability被夸大了事实上根据我的经验AMD和NV的OpenCL实现组合行为是有差异的并且有些十分隐蔽难于调试。而且同样的代码在AMD和NV是有性能差异的有时候差异非常大为了一致的性能不得不写两套代码。如果有更多的vendor呢 3.OpenCL作为开放的标准完全依赖于厂商的实现不同厂商支持标准不同。如果NV放弃支持OpenCL那它还是通用的开放的标准吗 4. 即使NV倒闭会有OpenCUDA出现的。 现在AMD在推新的HSA其IL类似NV的PTX不知是何用意。Java会在未来JAVA 9? 2015支持NV/AMD GPU 加速。OpenCL的美好只存在于未来问题是这个未来有多远。 CUDA助力OpenCL GPU并行计算无处不在
通用计算新锐OpenCL CUDA来助阵 通用计算新锐OpenCL CUDA来助阵 GPU经过多年的发展从功能单一的3D计算逐步扩充了视频解码、通用计算等而且值得一提的是通用计算这个目前最璀璨的技术新星被科研单位及个人消费者普遍关注。 众所周知NVIDIA是GPU的通用计算技术先驱者它的CUDA架构产品深入人心。而在通用计算的API层面不止NVIDIA一家独秀Apple苹果公司主导的OpenCL也赢得了业界同行的大力支持当然NVIDIA也是OpenCL的核心成员之一。 2008年12月9日2008亚洲SIGGRAPH大会上全球视觉计算技术的行业领袖NVIDIA公司于今日正式宣布完全支持Khronos Group新近发布的OpenCL 1.0技术规范。OpenCL开放式计算语言是一种全新计算应用程序接口 (API)它让开发人员能够利用GPU内部巨大的并行计算动力。OpenCL的加入是GPU革命史上又一重大的里程碑为NVIDIA开发人员提供了另一个功能强大的编程选择。 Khronos Group是一个非营利的会员组织的行业协会Khronos 致力于创造免授权费用的应用程序接口 API 及相关标准生态系统并将其标准及相关技术应用于包括手持设备控制台以及嵌入式系统中的高级动态多媒体应用以提供用以将这些多媒体技术转化为现实生产力的基础。例如OpenGL ES、OpenKODE、OpenMAX、OpenVG、OpenSL ES、COLLADA等当然还有今天的主角OpenCL。 OpenCL一经提出就得到了全球顶级软硬件厂商的广泛支持尤其是已经走在通用计算前沿的NVIDIA公司更是大力支持NVIDIA公司嵌入式内容副总裁Neil Trevett就在Khronos机构中担任OpenCL工作组主席一职。 Neil Trevett先生表示“OpenCL技术规范是NVIDIA等行业领袖意识到这一机遇之后所取得的成果。凭借其开放式、跨平台的标准该技术规范将在整个异构并行计算市场中得到更多的认可。NVIDIA将继续活跃于OpenCL工作组以便推动该技术规范的发展并在所有NVIDIA平台中提供对OpenCL的支持从而为开发人员提供另外一种方式来利用我们GPU中的超强计算动力。” OpenCL是什么与CUDA关系如何 OpenCL是什么与CUDA关系如何 OpenCL实际上是针对异构系统进行并行编程的一个全新API简单来说OpenCL它可以利用GPU进行一些并行计算方面的工作。这是API应用程序的编程接口图形里面也有很多API比如OpenGL、DirectX是针对图形的OpenCL是针对并行计算的API。 OpenCL开发人员可以利用GPU和CPU的计算能力把GPU和CPU异构的系统运用在很多并行计算的领域里面。Khronos就是这样一个组织它由很多厂商组成并有非常多的成员这个工作组同时也是OpenCL的一个协调机构也就是Khronos这个工作组它来负责制定OpenCL的规格、架构等等各方面。业界最主要的和图形或者和计算相关的厂商都是Khronos的成员。 值得一提的是OpenCL得到Khronos组织里很多成员的支持同时也验证大家对开发GPU并行计算的能力这个需求越来越多也正是因为有这个需求成就了OpenCL。 那么NVIDIA和OpenCL是一个怎样的关系呢 实际上OpenCL对于业界来说是非常重要也是非常好的一个标准它的出现令业界拥有了一个共同的标准利用GPU的强大计算能力然后应用在图形以外各种各样的并行计算中。其中NVIDIA一直也参与OpenCL的工作正如前文所提及的NVIDIA副总裁Neil Trevett先生他现在就任OpenCL工作组的主席引导很多OpenCL的开发当然这个组织中还有很多其他开发公司。 OpenCL最早是Apple公司提出来的从OpenCL一开始NVIDIA就和Apple公司进行非常紧密的合作。OpenCL开发的过程中它的技术平台都是NVIDIA的GPU实际上OpenCL是基于NVIDIA GPU的平台进行开发的。另外OpenCL的第一次演示也是在NVIDIA的GPU上演示的可以说到目前为止NVIDIA GPU几乎是进行OpenCL程序的唯一的平台。 另外还有非常重要的一点对于Apple公司来说他们是把GPU计算当成一种未来的趋势他们非常重视OpenCL他们将会在新一代的产品里面选择最适合于OpenCL运行的平台。所以说他们新一代苹果的笔记本电脑全都采用了NVIDIA的平台不管是MacBook Pro、Mac OSX都是采用了NVIDIA公司的平台。实际上这也是从另外一个方面证明NVIDIA的GPU对于OpenCL的支持到目前为止是最好的一个硬件的平台。 彻底了解CUDA架构/OpenCL存在意义 彻底了解CUDA架构/OpenCL存在意义 CUDA是什么很多人认为它是一个由NVIDIA设计的一种新软件或者新API不过笔者在此要告诉大家CUDA是一种硬件架构也就是说目前NVIDIA的GeForce产品全都基于CUDA架构设计。CUDA架构最主要的包含两个方面一个是ISA指令集架构第二硬件计算引擎。实际上它就是硬件和指令集这两个方面是CUDA的架构。 NVIDIA GPU的架构就是CUDA的架构举例来说你可以把它看成是跟Intel的X86或者IBM的Cell他们都是CPU架构而CUDA架构是基于GPU的架构。 CUDA的GPU架构和CPU架构很类似比如X86是包含一套指令集和执行X86各种各样的CPU而对于CUDA也是一样的NVIDIA有一套指令集ISA还有各种各样执行指令集和各种各样的硬件引擎。另外CUDA到目前为止它包含了一个C语言的编译器就是在CUDA上面的C语言CUDA这个架构还可以支持其他的API包括OpenCL或者DirectX同时以后NVIDIA的CPU还支持其他语言包括Fortran、Java、Python等各种各样的语言可以说这种架构是原生的专门为计算接口而建造的这样的一个架构。CUDA硬件架构包括指令集都是非常适合于并行计算为异构计算而设计的一整套架构。 也许会有人疑问既然有了CUDA为何还需要OpenCL或者类似OpenCL的API呢 简单来说对于编程人员他可以选择不同的方式来进行编程他们可以选择OpenCL API编程也可以选择C for CUDA语言来编程。 而API和语言的编程、开发存在着本质的不同API是一个编程接口它的核心是函数库和应用程序开发的一个硬件接口。用API来编程的话它有一个好处那就是可以访问比较低层次的硬件资源但这样的最大化控制硬件资源带来了很多弊端例如在内存管理上就必须靠编程人员手动控制这就需要编程人员要拥有高超的编程技术和深厚的经验积累。而就C for CUDA来说编程人员在利用C for CUDA语言来编程的时候无需考虑过多与自身编程目的以外的因素。再拿前文提及的内存管理来说C for CUDA使用Runtime进行管理。 不过不管OpenCL还是C for CUDA语言来编程最终它都是需要通过一个驱动程序来变成一个PTX的代码PTX相当于CUDA的指令集来进行执行然后交给图形处理其或者交给硬件来进行执行。这两个最终达到都是使用PTX或者在我们GPU上进行执行。 简单来说如果你想获得更多的对硬件上的控制权你可以使用API来进行变成。而如果你对API不是十分了解或者说无法很好的掌控API编程这时你可以用CUDA C语言来编程。二者是两种不同编程的方式它们有相同点也有不同点但是有一点OpenCL和CUDA C语言进行开发的时候在并行计算方面它们的概念十分接近这也就奠定了程序之间的相互移植会比较容易。 对C语言编程了解的读者应该知道C语言利用的驱动程序就是API也许谈及这个很抽象实际上C for CUDA就是一种C语言的扩展而针对扩展的主要方面就是并行运算编程这些是通过C的扩展来获得。基本上认为CUDA的程序也是一种标准的C语言的程序然后你使用一些关键字然后来对并行这方面计算然后做一些区分。C语言最终编译会成为PTX的代码然后在GPU上执行。 OpenCL是一个API就是应用程序的编程接口OpenCL和OpenGL很像这种API你可以调用API里面的函数库通过程序开发调用各种各样的函数实现各种各样的功能。对于API来说一般它对硬件设备有比较完整的访问权你可以访问硬件设备可以对内存进行管理最后OpenCL通过编译和驱动程序可以生成PTX代码在GPU上进行执行。 OpenCL未来发展目标激进 OpenCL未来发展目标激进 通过OpenCL的路线图我们能够看出目前OpenCL还处于Alpha版本但NVIDIA的产品已经可以非常好的支持OpenCL。明年第一季度Khronos Group会发布OpenCL Beta版本而OpenCL1.0正式版也可能在明年正式推出。不过OpenCL最开始可能出现在Mac OS上不过随着以后的逐渐扩展例如Windows或者Linux等系统都将会得以支持。 对于CUDA C语言NVIDIA一直不断地对其进行更深层次的开发同时也不断的有新版本的出现到目前为止已经是CUDA 2.0。CUDA C语言NVIDIA研发已经超过5年时间基本上从2003年左右就开始开发这个语言开发。 并且到目前为止开发人员的数量已经是超过25000人应用程序超过100个特别在科学计算的领域里CUDA的应用极为广泛几乎涉及到各种各样的HPC高性能计算的领域。甚至现在HPC进入排行榜前100的高性能计算机里面也有使用NVIDIA基于CUDA的Tesla系统它的开发语言正是使用了CUDA C语言。目前GPU集群组成的高性能计算机集群数量已经达到了30个在中国也有相关产品。 CUDA C语言是一个跨操作系统的开发工具现在支持Windows、Linux、Mac OS几乎把控了目前最主流的操作系统。NVIDIA的CUDA C语言提供了很多的库这些库主要包括FFT、BLAS等它们可以提供各种各样现成的代码让编程人员使用。 也许有这样的开发者他可能对CUDA并不熟悉但他又想要使用CUDA的高性能计算该怎么办很简单他就把现有的应用软件例如FFT使用CUDA的FFT的代码直接替换就这样他便可以获得20倍以上的性能提升。也就是说这些库是CUDA C语言非常重要的资源可以让大家缩短研发周期、体验CUDA带来的高性能例如一些数学软件像Matlab、Mathematica、LabView都有CUDA的插件可以使用C语言的插件让他更容易的利用CUDA。 CUDA/OpenCL共进退 前途无量 CUDA/OpenCL共进退 前途无量 这是NVIDIA的CUDA C语言路线图现在CUDA正处于2.0版本到今年年底NVIDIA可能会推出CUDA 2.1版本到明年会有CUDA 2.2、CUDA2.3版本最后迎来CUDA 3.0。随着CUDA版本的升级它的功能也在不断地升级比如最早的CUDA只能支持单精度的浮点计算现在可以支持双精度还可以支持各种各样的库。 总结的来说OpenCL无论对开发人员还是业界人员、消费者来说都是一个非常好的API一个应用程序的接口。它可以使开发者更容易的开发出跨平台GPU计算程序并将GPU强大的计算能力利用到各种各样应用计算中。 对于NVIDIA来说现在CUDA架构上除了C语言以外现在新增加了OpenCL或者DX11这样的API对于开发人员来说也是提供了一种更好的GPU计算的开发环境。因为对API很熟悉的开发人员他们肯定会很高兴的看到OpenCL或者新的API的加入对于这些人来说他们很容易利用这种API进行各种各样GPU计算程序开发。对于NVIDIA来说还会继续对C语言包括其他语言的支持实际上对NVIDIA CUDA C语言来说Runtime C仍是目前唯一的语言环境。今后除了C语言外NVIDIA还会推出更多的CUDA语言这包括Fortran还会有Java等。 性能与功能兼备 OpenCL和CUDA全解释 Khronos组织最近规范了OpenCL 1.0 OpenCL实际上是针对异构系统进行并行编程的一个全新的API简单来说OpenCL它可以利用GPU然后进行一些并行计算这方面的工作这是API应用程序的编程接口图形里面也有很多API比如OpenGL那是针对图形的OpenCL是针对并行计算的API。OpenCL开发人员可以利用GPU和CPU的计算能力把GPU和CPU异构的系统运用在很多并行计算的领域里面。Khronos是一个组织有很多厂商组成有非常多的成员这个工作组同时也是OpenCL的一个协调机构也就是Khronos这个工作组他来负责制定OpenCL的规格、架构等等各方面。业界最主要的和图形或者和计算相关的厂商都是Khronos的成员。 OpenCL里面有很多的成员这就导致了有更多可以利用OpenCL来开发的程序、软件以及各种各样的应用。这从一方面肯定了OpenCL强大的聚合实力但一方面也预示着这将会形成一个旺盛的需求。实际上OpenCL对于业界来说是非常重要也是非常好的一个标准NVIDIA看准了这一标准利用GPU的强大计算能力应用在图形以外各种各样的并行计算方面NVIDIA一直在参与OpenCL的工作。NVIDIA副总裁Neil Trevett目前则是OpenCL工作组的主席引导很多OpenCL的开发。这个组织里面当然还有很多其他开发公司。NVIDIA公司不少员工都在参与这项工作。 OpenCL最早是Apple公司提出来的由于OpenCL一开始NVIDIA就和Apple公司进行非常紧密的合作。所以目前OpenCL也完全是基于NVIDIA GPU的平台进行开发的。Apple十分重视OpenCL不管是MacBook Pro、Mac OSX等新产品采用了NVIDIA的平台。目前OpenCL路线图目前还是属于Alpha版本明年第一季度可能是Beta的版本09年时OpenCL1.0可能正式推出OpenCL最开始可能出现在Mac OS上以后逐渐的扩展到其他的操作系统像Windows或者Linux。 NVIDIA一直还是不断地对这个语言进行更深层次的开发到目前为止已经是CUDA 2.0。开发人员的数量已经超过25000个应用程序数量也已经超过100个特别是很多的科学计算的领域等几乎涉及到各种各样的HPC高性能计算的领域都有CUDA的身影出现甚至现在HPC进入排行榜前100的高性能计算机里面也有使用Tesla系统Tesla系统就是一个GPU集群计算典范现在支持Windows、Linux、Mac OS几乎最主流的操作系统NVIDIA都能支持。CUDA最主要的包含两个方面一个是ISA指令集架构第二硬件计算引擎实际上是硬件和指令集这两个方面是CUDA的架构。从这张图片可以看到当GPU变身成CUDA运算时候NVIDIA的GPU的架构就全是CUDA的架构你可以把它看成是跟X86或者cell他们都是架构这个也是CUDA架构但是是基于GPU的架构。这个和CPU的架构很类似比如X86是包含一套指令集和执行X86各种各样的CPU对于CUDA也是一样的NVIDIA自有一套指令集ISA还有各种各样执行指令集各种各样的硬件引擎。另外CUDA到目前为止它包含了一个C语言的编译器就是在CUDA上面的C语言CUDA这个架构还可以支持其他的API包括OpenCL或者DirectX同时以后NVIDIA还有其他的语言包括Fortran、Java、Python等各种各样的语言可以说这种架构是原生的专门为计算接口而建造的这样的一个架构NVIDIA的硬件架构包括指令集都是非常适合于这种并行计算这种异构计算而设计的一整套的架构。 CUDA和OpenCL的关系是不冲突的OpenCL是一个API在第一个级别CUDA这个架构是更高一个级别但是在这个架构上不管是OpenCL还是DX11这样的API还是像C语言、Fortran、DX11计算NVIDIA都可以支持。作为程序开发员来讲一般他只懂这些语言或者API他可以选择我用什么样的语言来开发我的程序不管他选择什么语言他希望调用GPU的计算能在这个架构上都可以用CUDA来编程对于编程人员来讲这个很容易他不需要对OpenCL的架构有非常深的了解他就可以做到。这个方面上面实际上是一些开发工具这个道理和CPU的编程实际上还是很类似的比如你有X86的指令集又有X86各种各样的CPU然后你只需要对这个指令集编辑X86架构上有各种各样的开发工具当然也有C语言Fortran语言有Python语言或者是其他的像Java或者以前的Pascal语言你用不同的语言进行开发最后你的执行还是在X86的架构上执行。NVIDIA的CUDA也是一样有CUDA这一套指令集有这样的硬件也就是说有不同的途径来进行开发你可以用OpenCL或者DirectX这样的API来进行计算的开发也可以用C语言或者Fortran或者Java开发这个道理是一模一样的可以做这方面的类比。OpenCL和C语言的一些异同点。对于编程人员来说他可以选择不同的东西来进行编程就像之前说的可以选择OpenCL编程也可以选择CUDA上面的C语言来编程或者API的语言来编程。API和C语言进行开发是有一些不同的API是一个编程接口它的核心是函数库和应用程序开发的一个硬件接口对于API来编程的话它有一个好处那就是可以访问比较低层次的硬件但是他也有一点就是很多的东西特别是像内存的管理这个是需要程序员自己来进行管理的。而C语言相对来说NVIDIA在利用CUDA C语言来编程的时候很多东西是由开发环境来进行管理的比如内存他是用runtime进行管理的runtime实际上就是运行时的一些支持程序来进行这方面的管理。不管OpenCL或者CUDA C语言来编程最终它都是需要通过一个驱动程序来变成一个PTX的代码PTX相当于CUDA的指令集来进行执行然后交给图形处理其或者交给硬件来进行执行。这两个最终达到都是使用PTX或者在GPU上进行执行。基本上大家可以理解为如果你想获得更多的对硬件上的控制权的话你可以使用API来进行编程比如我是一个科学家我对API不是太了解你也可以用CUDA C语言来编程这是两种不同编程的方式他们有他们相同点和不同点。但是有一点OpenCL和CUDA C语言进行开发的时候在并行计算这块他们的概念是差不多的这两种程序在程序上是有很大的相似度所以程序之间的相互移植相对来说也是比较容易。 大家如果使用了C语言的话都知道C语言使用驱动程序就是API实际上是一种抽象这个抽象主要是指和硬件相关的抽象。实际上CUDA C语言是一种C语言的扩展这扩展的一部分主要是进行并行运算编程的一方面这些是通过C的扩展来获得的。基本上认为CUDA的程序也是一种标准的C语言的程序然后你使用一些关键字然后来对并行这方面计算然后做一些区分。C语言最终编译会成为PTX的代码然后在GPU上执行。OpenCL是一个API就是应用程序的编程接口OpenCL和OpenGL很像这种API的话你可以调用API里面通过程序开发调用各种各样的函数实现各种各样的功能。对于API来说一般它对硬件设备有比较完整的访问权你可以访问硬件的设备可以对内存进行管理这是由开发人员通过编程来做的这些事情。最后完了OpenCL通过编译、通过驱动程序可以生成PTX代码在GPU上进行执行。邓老师讲的这个目的是如果你作为编程人员要利用GPU的计算能力开发你的应用的时候有两种模式——根据你不同的需求如果你需要对硬件有更多的控制你可以通过OpenCL来编你的API和你的程序那么它可以在CPU上运行如果你不需要这个同时对硬件有控制权而且你又不太懂API这些东西只要用C就可以编程了或者CUDA来编程编完程序以后也可以在CPU的硬件上跑这个东西不需要在CUDA支持GPU上跑。讲这个的目的是不同的编程模式可以选择不同的方法就是用OpenCL还是用CUDA C语言。 OpenCL和OpenGL在很多方面都很类似实际上他们也是一个共同的组织来进行管理的。对OpenGL图形开发比较熟悉的人他使用OpenCL计算这方面的开发他们就会非常熟悉它里面所涉及的很多方面这是OpenCL的一个非常明显的特点。如果你对图形编程很熟悉的话使用DX11编程可能比较容易这是API的一个好处。但是对于大部分的科学家来说可能对APIOpenCL这种东西可能完全不熟悉他需要的是我就像在CPU上编程一样对CPU的计算编程他可以使用CUDA C语言在CUDA C语言里面把CPU看成专门做计算的协处理器来进行编程的。这是两个之间不同的模式。
总结 OpenCL不管对开发人员还是业界人员来说还是消费者来说都是一个非常好的API一个应用程序的接口。它可以使得开发者很容易的开发出跨平台的GPU计算的程序充分利用GPU强大的计算能力然后应用在各种各样计算的方面。 对于NVIDIA来说在CUDA的架构上除了C语言以外现在新增加了OpenCL或者DX11这样的API对于开发人员来说也提供了一种更多的GPU计算的开发环境的一种选择。他们如果对API很熟悉的一批人他们肯定会很高兴的看到OpenCL或者新的API的加入对于这些人来说他们很容易利用这种计算的API然后开发各种各样GPU计算的程序。对于NVIDIA来说还会继续对C语言包括其他语言的支持实际上对NVIDIA CUDA C语言来说目前还是唯一的针对GPU的runtime C的语言环境runtime C的语言环境意思是GPU直接执行这个C语言。 现在已经有非常多的用户已经在使用CUDA C语言刚才介绍有25000名这样的开发者应用程序也有很多而且这种应用程序每一刻都在不断地增加数量。CUDA C语言还会进一步的发展就像刚才所说的还会有新的版本推出而且会和像OpenCL和DX11这种计算API会共存今后也是这样。除了C语言以外NVIDIA还会推出更多的其他CUDA的语言包括Fortran还会有Java等。 性能与功能兼备 OpenCL和CUDA全解释 编者按Apple是OpenCL创始人NVIDIA支持DirectX以及OpenGL这会形成一个对立还是和谐的合作从Apple最新的Macbook中可以看到两者已经很好地融合在一起。究竟这种关系是怎样融合起来的今天我们籍着这个话题为你一一解答 最近NVIDIA的CUDA技术成功地让GPU分担了CPU任务成功地让GPU拥有了CPU计算能力随着CUDA以及GeForce深入到PC更多的应用将会被提出。Apple最新产品Macbook也应用了NVIDIA的MCP7A芯片组以实现多媒体更高速运算。众所周知NVIDIA是DirectX以及OpenGL的忠实支持者以及这两个规范下的重要拥护者而Apple则是OpenCL的重要拥护者。究竟他们之间是一种怎样的关系下面是CUDA应用的各种例子 《GPU新里程 NV物理加速Demo体验测》 《发自内在的诱惑 9400M游戏及PS体验》 《开始说画 讯景9600GSO体验NV新技术》 在这些精彩的应用背后究竟隐藏着一个怎样的开发项目NVIDIA未来对于CUDA以及其项目有无进一步的消息究竟CUDA与CopnCL是怎样的一个融合体 关于OpenCL与CUDA的一些问题
问 AMD是OpenCL里面的吗
答 对。OpenCL是包含很多家公司的主要业界的公司都包含在里面包括IBM、戴尔这些公司包括HP。OpenCL是一个非常开放的行业组织所有的公司都在里面他们和NVIDIA是没有冲突的NVIDIA也是很积极地参与技术。 Stream基本上还是基于一种传统CPU的一种方式AMD当然他会说他会有CALCAL实际上是套指令集可以用汇编语言的方式来开发软件但是汇编方式开发软件的话对搞计算的人来说不大现实你要让他用汇编语言来说的话可能确实是一个折磨。还有一个Brook这个是斯坦福大学开发的它是类似于C的东西他是把底层GPGPU的计算方式类似于C的这种语言他不是C语言而是类C语言语法和C语言比较类似他内部做的还是使用顶点这样的数据。 另外Stream的方式还有一个很大的问题他主要是基于本地的板载内存板载内存存入数据然后换算完了再写到板载内存。这样对GPU非常强大的计算能力来说带宽是一个非常大的障碍你想想每秒种进行1P的数字计算的话你需要多少的带宽32位浮点的话这是4个字节如果你1P的话NVIDIA把他承加这部分也算上就相当于再乘以2至少需要每秒2P的吞吐量然后才能够满足它板载内存的话每秒需要几十P这实际上是Stream的方式从效率上是一个比较低的计算。而对于开发者来说也会碰到很多的问题对于NVIDIA来说支持C这是真正GPU上运行的C语言。C语言有一个很重要的特点需要有存储体系对于NVIDIA GPU来说是有存储体系的内部有share memory然后大部分的数据编译器会把大部分的数据尽可能的让它在share memory上进行计算share memory带宽非常高因为它在芯片内部它的速度接近于寄存器的速度他在share memory上跑的话然后再把这个数据再输出这样的话他真正地可以利用到绝大部分的计算机。实际上CUDA的效率在应用软件上可以看出来像传统的高性能计算领域的话NVIDIA的峰值速度比如像G80是300多G浮点计算能力效率是非常高的。这也是为什么这么多人在使用CUDA C语言来进行开发这是一个很重要的原因。 实际上你说要使用Stream来开发的应用软件我知道的只有一个是foldinghomeATI比NVIDIA进入早两年但是NVIDIA进入以后使用CUDA的语言来写foldinghome客户端的软件性能立刻比它高好几倍foldinghome你可以看到一个很特别的——他原来一个版本是可以跑在它上一代的架构应该是3850或者3870的GPU上但是如果它4850和4870出来以后从原点上来说他的计算能力比3850、3870要高2.5倍但是它的foldinghome的性能反而下降了原来他是200左右就是3870他是每天可以模拟200万秒的性能等级4870出来以后每天只能模拟170万秒。这是为什么这个软件不是使用Brook开发的直接使用汇编的方式开发的这样的话你需要对每一个新的GPU你都要对他进行编程每个架构都需要对它重新进行编程然后才可以得到一个最好的效率。中间有很长时间4870的性能都比3870在foldinghome上低后来他确实重新编程了过了好几个月有些新的版本出来了4870比3870高了。这就是说使用汇编方式的话会带来一个非常大的问题你任何一个应用软件里面都需要重新优化、重新编程。 汇编语言实际上就是机器码来进行编程这个是属于体力比较好的人才能干的了就是记忆也比较好的人才能干的了。高级语言相对来说确实要简单太多了就像刚才我说的有些人甚至对编程都不是很熟悉的他只要基本上知道源代码里面哪些代码是在做什么东西然后他就可以使用CUDA他不见得核心部分计算部分一定要自己编就像刚才说的NVIDIA有很多现成的库都是直接写好的函数这个函数把原来的函数替代掉就可以取得很好的性能。NVIDIA提供很多函数库这些函数库有一些是NVIDIA公司开发的像FFT、线性代数或者是快速傅里叶变换等但是有一些库是第三方帮NVIDIA开发的。NVIDIA实际上形成一个非常好的环境不断的有人针对CUDA来开发不管是应用软件或者是中间件这些中间件的话就包含各种各样的库。 对于CUDA来说说老实话和Stream相比的话几乎现在没什么可比性。确实对OpenCL、DirectX这样应用的话我相信在CUDA的架构上运行这个程序也比在Stream上运行程序好的多。因为NVIDIA使用的架构实际上你开发的时候NVIDIA就考虑到计算方面的应用。别的方案的话他有可能计算仅仅是图形的一个副产品。NVIDIA在开发的时候图形和计算这两部分几乎作为同等重要的两个都必须要满足的东西来做架构方面的设计。
问 OpenCL它的特点一个是开放OpenGL也很开放但是它的更新实际上非常、非常的慢OpenCL会有这种情况吗
答更新的话这是由一个专门的组织Khronos这个组织是有很多厂商来组成的每个厂商派一些人然后进来在组织里面然后每个公司互相协商然后制定各种各样的标准比如你推出一个新的版本要达到什么功能价格是怎么样的这些厂商来进行协商的厂商不太一样比如Microsoft我自己的东西我说的算这是完全不一样的。但是我相信OpenCL从目前来看的话它还是挺快的你看它现在的α版本明年一季度就进入β版本明年终第一版就会推出。
问 OpenGL和DirectX的关系一样我现在看到CUDA这边更新的频率明显比OpenCL快很多NVIDIA想把这个东西打造成一个GPU通用计算上的DirecX吗
答 我觉得应该是没有什么可比性的。刚刚强调DirectX是APICUDA不是APICUDA是一个硬件架构。刚才说的很多版本是CUDA C语言的版本这是CUDA C语言版本的路线图PPTC是语言CUDA是一个硬件架构GPU的架构。大家很容易会把C和CUDA融在一块儿不是这样的。CUDA架构类似与X86的架构。 实际上CUDA的C语言和OpenCL在并行计算这块它的一些观念都挺类似的包括它计算的数据组织方式都很类似。我相信NVIDIA在OpenCL的规格、开发方面联系一直是非常紧密的包括Apple也参与到其中。NVIDIA在这方面的很多经验也会得到一些体现。这两个程序之间的移植也是很容易的现在他使用CUDA C语言来写OpenCL出来以后如果他想移植OpenCL的程序相对来说不用全部重头写。
问 这个东西能否用在微软上
答 可以像OpenGL也不是微软的但是一样可以在微软上跑。
问 现在一般都是OpenCL
答 这个是程序模型的原因而且现在看起来在OpenCL在Vista下目前看起来没有太大的问题。原来大家都一直以为在Vista下面OpenCL不会有ICD实际上最后还是有ICD。这个东西的话现在OpenGL的软件我经常跑各种各样的专业软件很多在OpenGL下来以后都没有问题非常典型的OpenGL的应用程序跑地都挺好的。对于OpenCL或者OpenGL在微软下你可以看成是一个API这个API你是可以安装的比起微软自己原来带有的这个东西。这算是一种应用程序吧。
问 OpenCL看似更趋向于底层开发的API那相对于CUDA C来说OpenCL开发的程序执行效率会不会比CUDA C效率更高一些
答 完全不存在这个问题。说老实话效率高不高并不取决于你用的什么语言什么样的API更大程度上取决于你的代码你代码的优化程度是怎样的这个才是影响他代码的效率。
问 在OpenCL出来之前CUDA并行计算的优势已经很大了现在为什么还非要出一个OpenCL通过中间的东西来调用我感觉不是费一道事吗
答 OpenCL是一个跨平台的一种API这个API就像OpenGL一样它也是希望跨平台就是跨软件和硬件的APIOpenCL也是一样也是跨软件和硬件平台的API算是一个业界的标准不管是哪一家的原则上都可以支持OpenCL这个平台支持API。OpenCL不是NVIDIA推出来的OpenCL最开始是Apple提出来的但是NVIDIA确实是在这个标准的制定当中NVIDIA和Apple进行了很紧密的合作然后NVIDIA提供了很多支持包括硬件方面的支持。 模式不一样适用的人群也不一样。适用的人群也不一样我的理解是OpenCL和OpenGL他们在很多方面都非常的接近如果习惯使用OpenGL他如果用OpenCL来开发的话会感觉到轻车熟路就像是Computer shader也一样如果你是多媒体这方面的应用你又想使用图形又想使用GPU的来计算你用Computer shader来做很容易。但是对CUDA C来说它是用的人群是你能够使用C语言的人你都可以使用CUDA C来开发。使用人群完全是不一样的我想可能也有一些人虽然他也会C语言但是他还是希望用OpenCL来进行开发用他跨平台的特性。另外或者因为其他的一些原因包括他希望更多对硬件方面的控制然后他应用软件里面他可能希望对内存进行完全的控制这些用户的话他可能就会选用OpenCL来进行开发。这是两个不同的东西。我相信即使OpenCL出来以后CUDA C语言的使用者还是会越来越多的。
如果认为一种开发环境就可以取代其他的开发环境这个是不现实的。我举个例子来说在X86的架构上除了C语言以外还有Java、Fortran还有Pascal语言这些不可能互相取代的每种语言、每种API都有它使用的人群都有喜欢用的人。不同的语言、不同的API都会满足不同人群的。我觉得现在不是太多GPU计算和API语言不是太多现在还比较少NVIDIA还会不断地推出Java、Pascal或者C也会支持。如果以前没有Fortran那些老年人你要让他们学习C语言的话他们可能用Fortran用了几十年所以说让他开发的话他就比较痛苦现在有了Fortran他就可以直接用Fortran来写CUDA的程序说不同开发的途径有不同的用户这个是肯定的。 问 Havok现在怎么样了
答 Havok目前为止基本上只能使用CPU做物理计算而PhysX可以使用GPU、可以使用CPU也可以使用PPU来做物理方面的计算。当然这三种里面GPU的效率是最高的PPU的话估计现在也没什么人买GPU的效率是最高的而且他真的是可以带来10倍的性能提升这个性能提升在物理计算方面CPU和GPU相比的话性能提升10倍这是很常见的。而且PhysX它是可以作为一种最基本的平台还有很多第三方的中间件在上面基于PhysX。它实际上是利用PhysX的平台可以实现别的东西比如类似于人工智能。NVIDIA曾经给大家提供过一个美式橄榄球的游戏视频那个游戏真的是很酷但是我不能放给大家看那家公司要求NVIDIA只能提供视频给大家不能把游戏给大家看。真的是很酷非常大的场面那个场面里面每个观众都是一个不同的人有一个体育场9000人每个观众都是一个实际的人而且每个观众都是不太一样的。最主要他真正的把人的动作模拟的非常逼真你撞击撞到什么位置比如说你看到撞过来的你会下意识的做一些防守的动作。这些防守的动作人跑过来不同的方向动作也不一样的。比如你跌倒跌倒空中你可能会有下意识的支撑动作这些都会有以后游戏真的是会越来越逼真。NVIDIA所谓静态逼真和动态逼真就是这样。静态逼真像这个Crysis已经做的很逼真了但是真正动起来还是感觉比较僵硬就是整个场景里面不管人物的动作或者其他的运动都是事先设置好的就像放的录像一样而不是真正时时运算出来的。如果你使用大量物理计算的话真的可以获得很好的动态逼真度。如果要实时计算对硬件的要是就会很高如果持有CPU加速的话它的动态架构和NVIDIA的动态架构就有区别就会影响速度不一样。Havok用CPU来做物理计算它计算能力只能那么高你里面设置的物理效果相对来说就会很有限。而且CPU还有很重要的任务在人工智能或者场景管理的方面所以说使用GPU来做物理方面的加速是非常有效的。另外还有一点新的180驱动支持更多的GPU PhysX的配置你可以使用2片的显卡一片用来渲染一片用来做物理加速这是效率最高的一种方式。你用一片来做同时做渲染同时做物理计算这也是可以的和纯粹使用CPU做物理计算的话性能会高很多但是并行计算和渲染都是需要GPU的计算资源所以如果你用单独的一片显卡的话用来做物理加速的话性能又会得到一个很大的提升。大家最好可以测试一下这方面的效果。
问 比如像EA的100多款游戏它那个PhysX主要是用GPU运算还是CPU运算
答 只要是PhysX支持的游戏当然是GPU了。每一款游戏都可以采用不同的策略我相信这些游戏开发商的话也会采用不同的策略来处理这些事情当然他最好的效果肯定是使用GPU来做物理这方面的计算然后游戏里面放大量的物理效果这是最好的方式。当然也会有一些比较简化的模式就是说你没有NVIDIA的显卡然后给你提供一个比较简单物理效果比较少的模式。你还可以运行但是物理效果带来的各种各样的好处相对就比较少。这也是为什么跟大家强调的NVIDIA现在叫GeForce CUDA里面有很多的新功能PhysX是NVIDIA独有的这是别人没有办法做到的。你要想作为一个玩家你想享受更好的效果当然要在NVIDIA的显卡上运行才会有更好的效果。这个市场是开放的谁都可以去但是看谁能够跑在前面。
问 PhysX是跨平台的吗PhysX主要是用于游戏吗现在大部分的游戏基本上都是DirecX的。
答 游戏机也可以支持。OpenGL和DirecX的API是独立的现在OpenGL游戏确实越来越少所以大部分的游戏都是DirecX的。
问 PhysX不是跑在CPU上的
答 他是跑在GPU上。对于PhysX来说CPU也可以跑。 Badaboom 1.1介绍
Elemental公司会提供Badaboom 1.1版的。Badaboom输出形式还是在H.264这个已经成为业界的标准了。所有的手持设备都支持H.264因为H264编码效率确实比较高但是输入的形式情况千差万别输入现在有很多用不同的编码格式不同的封装格式。编码格式比如像WMA、Mpeg2、VC1等来自各种各样的渠道。还有封装格式比如像AVI、MKV等。不同的编码格式和封装格式现在Badaboom 1.1都可以支持这是一个方面。另外一个方面增加了更多的预设式的东西原来主要提供的是Apple的iphone、ipod或者iTV这种的支持新的提供更多的支持包括像youtubeyoutube这在国外用的很多另外还有Black Berry 9000的支持另外对微软播放器rom的支持。当然我还可以提供用户自定义的输出格式如果你想自定义分辨率自定义影片的细节你也可以自己定义包括码率等。第三非常重要的一点Badaboom 1.1可以支持Main Profile的H.264。过去都是Baseline的编码器Baseline和Main Profile最重要的区别是Main Profile可以使用基于二进制的可编程的双编码。它就是说在相同的码率下它的画质就更好。Main Profile一个非常重要的特点几乎所有的像蓝光的影片H.246的影片几乎都是采用Main Profile因为Main Profile可以提供更好的画质。现在像快速的编码器现在只有Elemental的Badaboom 1.1可以支持Main Profile。另外一个是支持全高清的输出1080P的输出现在的版本支持到720P以后可以支持到1018P的输出。还有一点是非常重要的可以支持多GPU。如果你有两个GPU的话可以获得性能翻番即是有两个GPU来进行编码。这是它的一些基本情况。 实际上对于Badaboom来说有一点非常重要Badaboom确实是非常容易使用而且真正是解放CPU的一种转码工具。Badaboom特别是在像H.264进行转码的时候H.264的影片比如蓝光的影片你把他转成ipod的影片CPU的占有率非常低所有的视频编码和解码全在GPU上完成。而且你在转码的过程中毕竟还需要二三十分钟转码过程中可以做别的没有任何问题。NVIDIA不像别的公司宣称什么同时利用CPU和GPU我都不太清楚他究竟用的是CPU还是GPU大家可以测试一下你看看它GPU的功耗有没有增加这是最简单的一个方法看看GPU有没有负载因为从原理上来说一个视频流除非你不怕数据在CPU和GPU来回倒要不然的话一个视频流很难把负载分到CPU和GPU上这两个数据处理的方式完全不一样的。 从另外一个角度就是画质这是非常重要的。编码的效率和画质是严重相关我可以这样说NVIDIA甚至可以做出一个可以损失一点画质然后我可以获得好几倍性能提升的编码因为H.264的编码太多了比如你帧间预测的时候你可以少拿几个关键帧或者你完全不要帧间预测只有帧队预测你也可以进行编码但是效果怎么样那是另外一回事。最终所有运动视频的编码最重要的一点是运动预测。如果多帧之间有运动的关系不用保存所有的图形你只要保存运动的适量就可以了所以大大降低文件的尺寸。如果你没有预测的话意味着每一张画面都要进行压缩你要达到比较小的分率的话压缩质量惨不忍睹。过去最早的Mpeg 2也可以使用CPU来做时时的压缩当然也可以买一个1万块钱的压缩卡1万块钱的压缩卡这和CPU压缩有什么区别CPU压缩的几乎没法看。我可以做到速度很高但是画质很差。NVIDIA希望在最好画质的情况下获得尽可能快的性能这也是Elemental这家公司想的。现在看起来Badaboom的画质和顶级的非常专业的编码工具得到的画质几乎是一样的但是它的性能比这些编码工具快很多。 我觉得好像预测然后“这个很快这个真的很快快了以后当然画质不太好这算是一个小缺点。”画质不太好这是一个大缺点不是一个小缺点NVIDIA转码的话是最好的画质当然是越快越好但是画质是绝对不能妥协的。就像3D游戏一样大家看一眼画质全世界都再说这个画质是不是有问题NVIDIA就是追求最好的画质然后尽可能提升性能。
问 像现在掌上设备不支持Cabac解码吗
答 掌上设备我还不太清楚是不是所有的不支持但是大部分的都支持Baseline但是还有一个Main Profile编码方式还是有一个非常大的好处比如你要想把一个1080P的视频转成720P码率小一点然后放在网络上让大家下载。
问 有些人在国外破解的BD都是原版的下载做成720P的。
答 以后可以用Badaboom比如Badaboom它可以自己定义输出的包括码率可以自己定义还是很方便现有的还是用H.264。
问 像纯做编码这块像Badaboom做这个它不是有编码加速吗
答 在这个过程中其他公司再做的时候NVIDIA也不能透露但是这肯定是他们的第一步后续的还会有做。在BD方面相对来说比较的快一点因为你要做核心来重新编写的话这个可能开发的周期也比较长。而且CUDA看起来做滤镜这方面还是特别实用的现在CUDA程序还有好几种。 谁主沉浮 OpenCL与CUDA架构深入解析 前言
最近Khronos公布了OpenCLOpen Computing Language的第一个测试版本一经发布便在通用计算领域掀起来轩然大波OpenCL是由苹果公司发起业界众多著名厂商共同制作的面向异构系统通用目的并行编程的开放式、免费标准也是一个统一的编程环境。便于软件开发人员为高性能计算服务器、桌面计算系统、手持设备编写高效轻便的代码而且广泛适用于多核心处理器(CPU)、图形处理器(GPU)、Cell类型架构以及数字信号处理器(DSP)等其他并行处理器在游戏、娱乐、科研、医疗等各种领域都有广阔的发展前景。 那么OpenCL与NVIDIA的CUDA架构是什么关系是否是外界认为的竞争关系目前众多的通用计算标准中比如NVIDIA的CUDA、Khronos的OpenCL、AMD的StreamCAL、Brook、微软下一代的Computer shader等他们之间有什么异同看完这篇文章相信你就会有一个大概的了解……
OpenCL来了
Khronos是一个开发组织著名的OpenGL就是出自Khronos之手Khronos有很多厂商组成OpenCL工作组同时也是OpenCL的一个协调机构来负责制定OpenCL的规格、架构等等各方面。业界最主要的和图形或者和计算相关的厂商都是Khronos的成员。
OpenCL实际上是针对异构系统进行并行编程的一个全新的APIOpenCL可以利用GPU进行一些并行计算的工作。我们知道图形里面有很多API比如OpenGL是针对图形的而OpenCL则是针对并行计算的API。 OpenCL开发人员可以利用GPU和CPU的计算能力把GPU和CPU异构的系统运用在很多并行计算的领域。 OpenCL对于业界来说是非常重要也是非常好的一个标准这样业界有一个共同的标准可以利用GPU的强大计算能力然后应用在图形以外各种各样的并行计算方面。NVIDIA公司的副总裁Neil Trevett是OpenCL工作组的主席引导很多OpenCL的开发NVIDIA公司很多员工都在参与这项工作。
OpenCL最早由Apple公司提出的OpenCL发起NVIDIA就和Apple公司进行非常紧密的合作。OpenCL开发的过程中技术平台均为NVIDIA的GPU实际上OpenCL是基于NVIDIA GPU的平台进行开发的。另外OpenCL在大概两个多月以前进行了第一次演示也是运行在NVIDIA的GPU上。 对于Apple公司来说是把GPU计算当成一种未来的趋势他们非常重视OpenCL在新一代的产品里面选择了最适合于OpenCL运行的平台。所以新一代苹果的笔记本电脑全都采用了NVIDIA的平台不管是MacBook Pro还是MaBook。实际上这也是从另外一个方面证明NVIDIA的GPU对于OpenCL的支持。
OpenCL与CUDA并非敌对关系
很多人对什么是CUDA可能还有一些疑虑并没有搞清楚CUDA到底是什么。实际上CUDA最主要的包含两个方面一个是ISA指令集架构第二硬件计算引擎实际上是硬件和指令集。 也就是说我们可以把CUDA看做是与X86或者cell类似的架构但是是基于是GPU而不是传统的CPU。
这个其实很好理解把它和传统的和CPU的架构比较下相信就更容易理解传统X86是包含一套指令集和执行X86各种各样的CPU对于CUDA也是一样CUDA有一套指令集ISA还有执行指令集各种各样的硬件引擎。CUDA到目前为止包含了一个C语言的编译器当然CUDA架构还可以支持其他的API包括OpenCL或者DirectX同时CUDA还会有其他的语言包括Fortran、Java、Python等各种各样的语言可以说CUDA架构是原生的专门为计算接口而建造的这样的一个架构这种硬件架构包括指令集都是非常适合于这种并行计算为异构计算而设计的一整套的架构。 OpenCL与CUDA的关系上图很好的解释CUDA和OpenCL的关系他们并不是冲突关系而是包容关系。OpenCL是一个API在第一个级别CUDA架构是更高一个级别在这个架构上不管是OpenCL还是DX11这样的API还是像C语言、Fortran、DX11计算都可以支持。作为程序开发员来讲一般他们只懂这些语言或者API可以采用多种语言开发自己的程序不管他选择什么语言只要是希望调用GPU的计算能在这个架构上都可以用CUDA来编程。
CUDA编程的道理和CPU的编程很类似比如有了X86的指令集又有X86各种各样的CPU那么我们只需要对这个指令集编程即可。X86架构上有各种各样的开发工具当然也有C语言Fortran语言Python语言Java或者以前的Pascal语言不论你使用什么语言进行开发最后还是在X86的架构上执行。CUDA也是一样有了CUDA的指令集有了支持CUDA的硬件我们就可以采用不同的途径来进行开发比如可以采用OpenCL或者DirectX这样的API也可以用C语言或者Fortran或者Java开发最终都可以在CUDA架构上运行。 OpenCL与CUDA C语言的异同
前文已经说过CUDA架构与OpenCL是包容关系我们把他们放在同一级别进行讨论本来就是错误的与OpenCL在同一级别不是CUDA架构而是CUDA的C语言包也就是我们常说的CUDA版本比如CUDA 1.0、CUDA 2.0等等下面来介绍一下OpenCL和CUDA C语言的一些异同。 对于编程人员来说可以选择不同的开发环境来进行编程例如我们可以选择OpenCL编程也可以选择CUDA上面的C语言来编程或者API的语言来编程。API和C语言进行开发是有一些不同的API是一个编程接口它的核心是函数库和应用程序开发的一个硬件接口对于API来编程的话好处在于可以访问比较低层次的硬件单弊端也是显而易见的那就是很多程序特别是像内存的管理需要程序员自己来进行管理。而我们在利用CUDA C语言来编程的时候底层的硬件管理是由CUDA开发包来进行管理比如内存是用runtime进行管理runtime实际上就是运行时的一些支持程序来进行底层硬件的管理而不需要开发者考虑底层的硬件效率。不管OpenCL或者CUDA C语言来编程最终都是需要通过一个驱动程序来变成一个PTX的代码PTX相当于CUDA的指令集来进行执行然后交给图形处理其或者交给CPU来进行执行。
也就是说如果开发人员想获得更多的对硬件上的控制权的话可以使用API来进行编程而如果类似科学家如果对API不是太了解那么就可以用CUDA C语言来编程这是两种不同编程的方式他们有他们相同点和不同点。OpenCL和CUDA C语言进行开发的时候在并行计算上的概念很相似在程序上是也有很大的相似度所以程序之间的相互移植相对来说也比较容易。
当然我们也许会有疑问那就是OpenCL看似更趋向于底层开发的API那相对于CUDA C来说OpenCL开发的程序执行效率会不会比CUDA C效率更高一些实际上这个问题是不用担心的执行效率高不高并不取决于采用的什么语言什么样的API更大程度上取决于的代码的优化程度
不同编程模式解读
下面我们来深入解读一下CUDA与C语言编程模式的异同。对C语言进行编程的人员都知道C语言使用驱动程序就是API实际上是一种抽象这个抽象主要是指和硬件相关的抽象。实际上CUDA C语言是一种C语言的扩展这扩展的一部分主要是进行并行运算编程的方面这些是通过C的扩展来获得的。基本上认为CUDA的程序也是一种标准的C语言的程序然后使用一些关键字然后来对并行这方面计算最后做一些区分。C语言最终编译会成为PTX的代码然后在GPU上执行。OpenCL是一个API就是应用程序的编程接口OpenCL和OpenGL很像这种API可以调用API最底层的数据通过程序开发调用各种各样的函数实现各种各样的功能。对于API来说一般它对硬件设备有比较完整的访问权以访问硬件的设备可以对内存进行管理这是由开发人员通过编程来做的这些事情。最后OpenCL通过编译、通过驱动程序可以生成PTX代码在GPU上进行执行。 编程人员要利用GPU的计算能力开发你的应用的时候有两种模式如果你需要对硬件有更多的控制可以通过OpenCL来编你的API和你的程序那么它可以在CPU上运行如果不需要底层的硬件管理同时对硬件有控制权而且又不太懂API这些程序只要用C语言就可以编程了或者CUDA C语言来编程编完程序以后也可以在CPU的硬件上运行。
OpenCL和OpenGL在很多方面都很类似他们也是一个共同的组织Khronos来进行管理的。对OpenGL图形开发比较熟悉的人使用OpenCL计算这方面的开发他们就会非常熟悉它里面所涉及的很多方面这是OpenCL的一个非常明显的特点。如果你对图形编程很熟悉的话就像是Computer shader如果是多媒体等方面的应用又想使用图形又想使用GPU的来计算用Computer shader的DX11编程可能比较容易。但是对于大部分的科学家来说可能对APIOpenCL可能完全不熟悉他需要的就像在CPU上编程一样对CPU的计算编程可以使用CUDA C语言在CUDA C语言里面我们把CPU看成专门做计算的协处理器来进行编程这是两个之间不同的模式。
如果认为一种开发环境就可以取代其他的开发环境这个是不现实的。举个例子在X86的架构上除了C语言以外还有Java、Fortran还有Pascal语言这些不可能互相取代的每种语言、每种API都有它使用的人群。不同的语言、不同的API都会满足不同人群的。GPU计算和API语言不是太多目前还比较少NVIDIA还会不断地推出Java、Pascal或者C也会支持。如果以前没有Fortran那些“老古董”程序员要让他们学习C语言的话基本不可行他们可能用Fortran用了几十年让他们使用C语言开发的话他就比较痛苦所以OpenCL与CUDA C语言并不存在谁将取代谁的问题。
CUDA与ATI Stream异同
我们知道ATI方面也有自己的通用计算编程接口叫做Stream。那么他与CUDA架构又有什么不同呢
Stream基本上还是基于一种传统CPU的一种方式Stream主要包括CAL与Brook。CAL是一套指令集可以用汇编语言的方式来开发软件然而我们汇编方式开发软件的话对搞计算的人来说不大现实让他们用汇编语言来说的话可能确实是一个折磨。Brook是斯坦福大学开发的它是类似于C语言的东西是把底层GPGPU的计算方式类似于C的这种语言这里要说明的是Brook不是C语言而是类C语言语法和C语言比较类似但内部做的还是使用顶点这样的数据。 Stream视频编码另外Stream的方式还有一个很大的问题那就是主要是基于本地的板载内存存入数据换算完成再写到板载内存。这样对GPU非常强大的计算能力来说带宽是一个非常大的障碍。例如每秒种进行1P的数字计算的话需要多少的带宽32位浮点4个字节如果是1P的话我们把乘加这部分也计算在内就相当于再乘以2至少需要每秒2P的吞吐量然后才能够满足板载内存每秒需要几十P。可见Stream的方式从效率上是一个比较低的计算而对于开发者来说也会碰到很多的问题。
对于CUDA来说我们支持的C语言是真正GPU上运行的C语言。C语言有一个很重要的特点需要有存储体系而NVIDIA GPU在设计之初就考虑到了这点内部有拥有share memory编译器会把大部分的数据尽可能的在share memory上进行计算。share memory带宽非常高因为它在芯片内部速度接近于寄存器。数据在share memory上运行然后再把这个数据再输出这样的话真正地可以利用到绝大部分的GPU性能。
汇编语言的弊端
基于Stream来开发的应用软件我们知道的只有一个是foldinghomeATI比NVIDIA进入早两年但是NVIDIA进入以后使用CUDA的语言来写foldinghome客户端的软件性能比ATI的GPU要高好几倍。foldinghome的一个版本可以运行在在上一代的架HD3报价参数 图片 评测850或者HD3870的GPU上但是HD4850和HD4870发布后从浮点性能上来说计算能力比HD3850、HD3870要高2.5倍但是foldinghome的性能反而下降了原来HD3870是200ns/day左右每天可以模拟200万秒的性能等级但是HD4870发布后每天只能模拟170ns/day这又是为什么其原因是软件不是使用Brook开发的而直接使用汇编的方式这就需要对每一个新的GPU进行编程每个架构都需要对它重新进行编程然后才可以得到一个最好的效率这就是说使用汇编方式的话会带来一个非常大的问题任何一个应用软件里面都需要重新优化、重新编程。 汇编语言实际上就是机械码来进行编程属于体力比较好、记忆也比较好的人才能胜任的工作。高级语言相对来说确实要简单太多了例如有些人甚至对编程都不是很熟悉他只要基本上知道源代码里面哪些代码是在做什么东西然后他就可以使用CUDA核心部分计算部分一定要自己编写因为CUDA有很多现成的库都是直接写好的函数这个函数把原来的函数替代掉就可以取得很好的性能。NVIDIA提供很多函数库这些函数库有一些是NVIDIA公司开发的像FFT、线性代数或者是快速傅里叶变换等但是有一些库是第三方帮NVIDIA开发的。CUDA实际上形成一个非常好的环境不断的有人针对CUDA来开发不管是应用软件或者是中间件这些中间件的话就包含各种各样的库。
因此对于CUDA来说与Stream相比几乎现在没什么可比性。对OpenCL、DirectX这样应用在CUDA的架构上运行程序也比在Stream上运行程序好的多。因为NVIDIA在使用的CUDA架构实际上字开发的时候就考虑到通用计算方面的应用。而ATI方面可能通用计算仅仅是图形的一个副产品。
OpenCL与CUDA C语言路线图
目前OpenCL路线图目前还是属于Alpha版本。明年第一季度可能是Beta的版本09年OpenCL1.0可能会正式推出OpenCL最早会出现在Mac OS上以后逐渐的扩展到其他的操作系统像Windows或者Linux。 OpenCL路线图下面来介绍一下CUDA C语言的情况NVIDIA一直还是不断地对这个语言进行更深层次的开发也不断的有新版本的出现到目前为止已经是CUDA 2.0。CUDA C语言研发已经超过5年时间NVIDIA从03年左右就开始开发C语言。到目前为止开发人员的数量已经是超过25000个而且现在应用程序已经超过100个特别是很多的科学计算的领域几乎涉及到各种各样的HPC高性能计算的领域都有CUDA的身影出现甚至现在HPC进入排行榜前100的高性能计算机里面也有使用NVIDIA的Tesla系统GPU集群就是配置成高性能计算机的集群数量已经达到了30个我国大陆也有采用。
CUDA C语言是一个跨操作系统的开发工具现在支持Windows、Linux、Mac OS几乎最主流的操作系统都支持SUN公司的Solaris系统也会支持。一些数学软件像Matlab、Mathematica、LabView都有CUDA的插件可以使用C语言的插件让他更容易的利用CUDA。 上面是C语言的路线图目前的版本是CUDA 2.0到今年年底CUDA 2.1会与我们见面到明年会有2.2、2.3版本到明年年底会到CUDA 3.0版本。随着CUDA版本的升级它的功能也在不断地升级比如最早的CUDA只能支持单精度的浮点计算现在可以支持双精度可以支持各种各样的库各种各样的功能也是越来越多。
并行计算时代已经到来……
OpenCL的面世不管对开发人员还是业界人员来说还是消费者来说都是一个非常好的API它可以使得开发者很容易的开发出跨平台的GPU计算的程序充分利用GPU强大的计算能力然后应用在各种各样计算的方面。 除了CUDA的架构上除了C语言以外现在新增加了OpenCL或者DX11这样的API对于开发人员来说也提供了一种更多的GPU计算的开发环境的一种选择。如果对API很熟悉的程序员肯定会很高兴的看到OpenCL或者新的API的加入对于这些人来说他们很容易利用这种计算的API然后开发各种各样GPU计算的程序。 NVIDIA也会继续对C语言包括其他语言的支持实对NVIDIA CUDA C语言来说目前还是唯一的针对GPU的runtime C的语言环境runtime C的语言环境是指GPU直接执行这个C语言。CUDA C语言还会进一步的发展不断会有新的版本推出。CUDA C语言会和OpenCL和DX11这种计算API会共存。 除了C语言以外NVIDIA还会推出更多的其他CUDA的语言包括Fortran还会有Java等。不管C语言还是Fortran与OpenCL、Computer shader这种API是一种长期共存的关系。 GPU通用计算时代已经到来了你已经准备好了吗…… OpenCL和CUDA的使用比较 OpenCL和CUDA虽然不是同一个平级的东西,但是也可以横向比较! 对OpenCL和CUDA的异同做比较: 指针遍历 OpenCL不支持CUDA那样的指针遍历方式, 你只能用下标方式间接实现指针遍历. 例子代码如下: // CUDA struct Node { Node* next; } n n-next; // OpenCL struct Node { unsigned int next; } n bufBase n; Kernel 程序异同 CUDA的代码最终编译成显卡上的二进制格式最后由cudart.dll(个人猜测)装载到GPU并且执行。OpenCL中运行时库中包含编译器 使用伪代码程序运行时即时编译和装载。这个类似JAVA, .net 程序道理也一样为了支持跨平台的兼容。kernel程序的语法也 有略微不同如下 [c-sharp] view plaincopyprint? __global__ void vectorAdd(const float * a, const float * b, float * c) { // CUDA int nIndex blockIdx.x * blockDim.x threadIdx.x; c[nIndex] a[nIndex] b[nIndex]; } [cpp] view plaincopyprint? __kernel void vectorAdd(__global const float * a, __global const float * b, __global float * c) { // OpenCL int nIndex get_global_id(0); c[nIndex] a[nIndex] b[nIndex]; } 可以看出大部分都相同。只是细节有差异 1CUDA 的kernel函数使用“__global__”申明而OpenCL的kernel函数使用“__kernel”作为申明。 2OpenCL的所有参数都有“__global”修饰符代表这个参数所指地址是在全局内存。 3众所周知CUDA采用threadIdx.{x|y|z}, blockIdx.{x|y|z}来获得当前线程的索引号而OpenCL 通过一个特定的get_global_id()函数来获得在kernel中的全局索引号。OpenCL中如果要获得在当前工作 组对等于CUDA中的block中的局部索引号可以使用get_local_id() Host代码的异同 把上面的kernel代码编译成“vectorAdd.cubin”CUDA调用方法如下 [cpp] view plaincopyprint? const unsigned int cnBlockSize 512; const unsigned int cnBlocks 3; const unsigned int cnDimension cnBlocks * cnBlockSize; CUdevice hDevice; CUcontext hContext; CUmodule hModule; CUfunction hFunction; // create CUDA device context cuInit(0); cuDeviceGet(hContext, 0); // pick first device cuCtxCreate(hContext, 0, hDevice)); cuModuleLoad(hModule, “vectorAdd.cubin”); cuModuleGetFunction(hFunction, hModule, vectorAdd); // allocate host vectors float * pA new float[cnDimension]; float * pB new float[cnDimension]; float * pC new float[cnDimension]; // initialize host memory randomInit(pA, cnDimension); randomInit(pB, cnDimension); // allocate memory on the device CUdeviceptr pDeviceMemA, pDeviceMemB, pDeviceMemC; cuMemAlloc(pDeviceMemA, cnDimension * sizeof(float)); cuMemAlloc(pDeviceMemB, cnDimension * sizeof(float)); cuMemAlloc(pDeviceMemC, cnDimension * sizeof(float)); // copy host vectors to device cuMemcpyHtoD(pDeviceMemA, pA, cnDimension * sizeof(float)); cuMemcpyHtoD(pDeviceMemB, pB, cnDimension * sizeof(float)); // setup parameter values cuFuncSetBlockShape(cuFunction, cnBlockSize, 1, 1); cuParamSeti(cuFunction, 0, pDeviceMemA); cuParamSeti(cuFunction, 4, pDeviceMemB); cuParamSeti(cuFunction, 8, pDeviceMemC); cuParamSetSize(cuFunction, 12); // execute kernel cuLaunchGrid(cuFunction, cnBlocks, 1); // copy the result from device back to host cuMemcpyDtoH((void *) pC, pDeviceMemC, cnDimension * sizeof(float)); delete[] pA; delete[] pB; delete[] pC; cuMemFree(pDeviceMemA); cuMemFree(pDeviceMemB); cuMemFree(pDeviceMemC); OpenCL的代码以文本方式存放在“sProgramSource”。 调用方式如下 [cpp] view plaincopyprint? const unsigned int cnBlockSize 512; const unsigned int cnBlocks 3; const unsigned int cnDimension cnBlocks * cnBlockSize; // create OpenCL device context cl_context hContext; hContext clCreateContextFromType(0, CL_DEVICE_TYPE_GPU, 0, 0, 0); // query all devices available to the context size_t nContextDescriptorSize; clGetContextInfo(hContext, CL_CONTEXT_DEVICES, 0, 0, nContextDescriptorSize); cl_device_id * aDevices malloc(nContextDescriptorSize); clGetContextInfo(hContext, CL_CONTEXT_DEVICES, nContextDescriptorSize, aDevices, 0); // create a command queue for first device the context reported cl_command_queue hCmdQueue; hCmdQueue clCreateCommandQueue(hContext, aDevices[0], 0, 0); // create compile program cl_program hProgram; hProgram clCreateProgramWithSource(hContext, 1, sProgramSource, 0, 0); clBuildProgram(hProgram, 0, 0, 0, 0, 0);// create kernel cl_kernel hKernel; hKernel clCreateKernel(hProgram, “vectorAdd”, 0); // allocate host vectors float * pA new float[cnDimension]; float * pB new float[cnDimension]; float * pC new float[cnDimension]; // initialize host memory randomInit(pA, cnDimension); randomInit(pB, cnDimension); // allocate device memory cl_mem hDeviceMemA, hDeviceMemB, hDeviceMemC; hDeviceMemA clCreateBuffer(hContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, cnDimension * sizeof(cl_float), pA, 0); hDeviceMemB clCreateBuffer(hContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, cnDimension * sizeof(cl_float), pA, 0); hDeviceMemC clCreateBuffer(hContext, CL_MEM_WRITE_ONLY, cnDimension * sizeof(cl_float), 0, 0); // setup parameter values clSetKernelArg(hKernel, 0, sizeof(cl_mem), (void *)hDeviceMemA); clSetKernelArg(hKernel, 1, sizeof(cl_mem), (void *)hDeviceMemB); clSetKernelArg(hKernel, 2, sizeof(cl_mem), (void *)hDeviceMemC); // execute kernel clEnqueueNDRangeKernel(hCmdQueue, hKernel, 1, 0, cnDimension, 0, 0, 0, 0); // copy results from device back to host clEnqueueReadBuffer(hContext, hDeviceMemC, CL_TRUE, 0, cnDimension * sizeof(cl_float), pC, 0, 0, 0); delete[] pA; delete[] pB; delete[] pC; clReleaseMemObj(hDeviceMemA); clReleaseMemObj(hDeviceMemB); clReleaseMemObj(hDeviceMemC); 初始化部分的异同 CUDA 在使用任何API之前必须调用cuInit(0)然后是获得当前系统的可用设备并获得Context。 cuInit(0); cuDeviceGet(hContext, 0); cuCtxCreate(hContext, 0, hDevice)); OpenCL不用全局的初始化直接指定设备获得句柄就可以了 cl_context hContext; hContext clCreateContextFromType(0, CL_DEVICE_TYPE_GPU, 0, 0, 0); 设备创建完毕后可以通过下面的方法获得设备信息和上下文 size_t nContextDescriptorSize; clGetContextInfo(hContext, CL_CONTEXT_DEVICES, 0, 0, nContextDescriptorSize); cl_device_id * aDevices malloc(nContextDescriptorSize); clGetContextInfo(hContext, CL_CONTEXT_DEVICES, nContextDescriptorSize, aDevices, 0); OpenCL introduces an additional concept: Command Queues. Commands launching kernels and reading or writing memory are always issued for a specific command queue. A command queue is created on a specific device in a context. The following code creates a command queue for the device and context created so far: cl_command_queue hCmdQueue; hCmdQueue clCreateCommandQueue(hContext, aDevices[0], 0, 0); With this the program has progressed to the point where data can be uploaded to the device’s memory and processed by launching compute kernels on the device. Kernel Creation CUDA kernel 以二进制格式存放与CUBIN文件中间其调用格式和DLL的用法比较类似先装载二进制库然后通过函数名查找 函数地址最后用将函数装载到GPU运行。示例代码如下 CUmodule hModule; cuModuleLoad(hModule, “vectorAdd.cubin”); cuModuleGetFunction(hFunction, hModule, vectorAdd); OpenCL 为了支持多平台所以不使用编译后的代码采用类似JAVA的方式装载文本格式的代码文件然后即时编译并运行。 需要注意的是OpenCL也提供API访问kernel的二进制程序前提是这个kernel已经被编译并且放在某个特定的缓存中了。 // 装载代码即时编译 cl_program hProgram; hProgram clCreateProgramWithSource(hContext, 1, “vectorAdd.c, 0, 0); clBuildProgram(hProgram, 0, 0, 0, 0, 0); // 获得kernel函数句柄 cl_kernel hKernel; hKernel clCreateKernel(hProgram, “vectorAdd”, 0); 设备内存分配 内存分配没有什么大区别OpenCL提供两组特殊的标志CL_MEM_READ_ONLY 和 CL_MEM_WRITE_ONLY 用来控制内存 的读写权限。另外一个标志比较有用CL_MEM_COPY_HOST_PTR 表示这个内存在主机分配但是GPU可以使用运行时会自动 将主机内存内容拷贝到GPU主机内存分配设备内存分配主机拷贝数据到设备3个步骤一气呵成。 // CUDA CUdeviceptr pDeviceMemA, pDeviceMemB, pDeviceMemC; cuMemAlloc(pDeviceMemA, cnDimension * sizeof(float)); cuMemAlloc(pDeviceMemB, cnDimension * sizeof(float)); cuMemAlloc(pDeviceMemC, cnDimension * sizeof(float)); cuMemcpyHtoD(pDeviceMemA, pA, cnDimension * sizeof(float)); cuMemcpyHtoD(pDeviceMemB, pB, cnDimension * sizeof(float)); // OpenCL hDeviceMemA clCreateBuffer(hContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, cnDimension * sizeof(cl_float), pA, 0); hDeviceMemB clCreateBuffer(hContext, CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR, cnDimension * sizeof(cl_float), pA, 0); hDeviceMemC clCreateBuffer(hContext, CL_MEM_WRITE_ONLY, cnDimension * sizeof(cl_float), 0, 0); Kernel Parameter Specification The next step in preparing the kernels for launch is to establish a mapping between the kernels’ parameters, essentially pointers to the three vectors A, B and C, to the three device memory regions, which were allocated in the previous section. Parameter setting in both APIs is a pretty low-level affair. It requires knowledge of the total number , order, and types of a given kernel’s parameters. The order and types of the parameters are used to determine a specific parameters offset inside the data block made up of all parameters. The offset in bytes for the n-th parameter is essentially the sum of the sizes of all (n-1) preceding parameters. Using the CUDA Driver API: In CUDA device pointers are represented as unsigned int and the CUDA Driver API has a dedicated method for setting that type. Here’s the code for setting the three parameters. Note how the offset is incrementally computed as the sum of the previous parameters’ sizes. cuParamSeti(cuFunction, 0, pDeviceMemA); cuParamSeti(cuFunction, 4, pDeviceMemB); cuParamSeti(cuFunction, 8, pDeviceMemC); cuParamSetSize(cuFunction, 12); Using OpenCL: In OpenCL parameter setting is done via a single function that takes a pointer to the location of the parameter to be set. clSetKernelArg(hKernel, 0, sizeof(cl_mem), (void *)hDeviceMemA); clSetKernelArg(hKernel, 1, sizeof(cl_mem), (void *)hDeviceMemB); clSetKernelArg(hKernel, 2, sizeof(cl_mem), (void *)hDeviceMemC); 携手OPENCL CUDA架构能否成惟一领袖 如果你是苹果的粉丝你一定会结识一个新朋友——NVIDIA如果你是NVIDIA的粉丝那你一定欣喜自己与万人迷“苹果”走的更近。在今年10月24日老乔津津有味的拿着一串数据来宣布旗下MacBook全系列产品以及今后相关众多Mac产品将采用NVIDIA相关方案苹果的粉丝被教父的Keynote所打动开始学习只有DIY发烧友才关注的一些技术名词他们在论坛里热情高涨讨论着“HybridSLI”、“CUDA”、“PureVideo”等等。 视觉革命与计算革命 苹果与NVIDIA本月又联手上演好戏本月苹果又向大家宣布的OPENCL标准它将在苹果明年初推出的新一代操作系统Snow Leopard上开始支持虽然我们看到支持的成员几乎涵盖了业内所有的知名企业但大家应该清楚在孕育OPENCL的过程中能够提供硬件平台的只有NVIDIA当然OPENCL标准工作组Khronos也首次将如此多的业内精英撮合到一起为了一个“并行计算”的新API。 大家为什么要支持OPENCL首先简单了解一下OPENCL和异构系统OPENCL英文全称Open Computing Language是基于异构系统并行编程的一个全新API命名上虽然与OPENGLOpen Graphics Library类似但方向上有很大不同但开发上又存在一些联系。在未来计算领域异构系统已经被公认为发展趋势即CPU与平行处理器的异构形式这也是未来Intel的发展方向。而我们目前能看到的能够提供异构系统并有成熟软件平台支持的仅有NVIDIA的CUDA架构。 在OPENCL标准组织中由于Intel和AMD的加入让大家对CUDA与OPENCL的关系产生很多疑问本来已经成熟的CUDA平台为什么还需要一个新API的加入Intel和AMD如何和OPENCL兼容OPENCL和CUDA谁更好这都是非常集中的问题。 我们来看看Khronos OPENCL组织是如何组成的。我们知道在OPENGL时期该标准也是由Khronos这样一个协调机构来运作的。而你会发现这次OPENCL由苹果作为代表发布也正是苹果选择NVIDIA战略中的重要一环与其说OPENCL是苹果发起的还不如说是NVIDIA策划的。而事实的确如此NVIDIA副总裁Neil Trevett正是现在Khronos OPENCL工作的主席OPENCL的标准的创立也是NVIDIA技术为基础的其中骨干员工也大半来自NVIDIA。其实这也不难理解要做一个异构并行计算的系统和标准API除了NVIDIA你还能找谁呢 而下面这张图也许是在此次OPENCL发布后较为常见的一个不过其中意义大家不一定能够完全领会我们在这里做一个说明。这是一张能够解释大家对OPENCL与CUDA关系疑问的重要“结构”图。NVIDIA第一次明确了两个重要名词“CUDA架构”与“CUDA C语言”。在以往理解中我们常默认CUDA就是指基于NVIDIA GPU或并行处理器的编译器、C语言、库和相关软件开发环境而硬件方面则是指所有支持CUDA的GPU。今天NVIDIA明确了包括硬件支持和软件环境一起我们称之为CUDA架构而常用的基于C语言的开发环境则是CUDA C或者CUDA FortranCUDA C以及今后的CUDA JAVA等等。而OPENCL也与之并列。 具体来说NVIDIA将CUDA架构看作是一个异构计算中并行计算部分的核心无论你使用怎样的高级语言或者是怎样的API的形式进行开发在硬件上均符合CUDA架构即NVIDIA GPU/并行处理器。这有点类似于传统CPU领域无论你使用何种高级语言、汇编语言、API进行软件开发但是都为X86架构服务。 解决了OPENCL与CUDA确切的说是与CUDA C的争论的问题大家仍心存疑问。NVIDIA在初期宣传CUDA C时其优势就在于高级语言对软件开发的便利和普遍性而OPENCL API级别的编程相对C语言来说又有什么优势和必要我们认为此时NVIDIA宣传CUDA与1年前初期阶段已经有所不同高级语言的支持是CUDA努力的主要方向而如何全方位的吸引开发人员的加入这对CUDA来说同样重要。 CUDA C擅长与服务的对象 OPENCL的特点与服务对象 我们从图中看到API和C语言在开发技术上有所不同NVIDIA表示对原有OPENGL开发人员来说一定会喜欢OPENCL因为其中有很多类似之处同时NVIDIA也表示由于都是基于CUDA架构所以OPENCL与CUDA C之间如果真需要迁移也是相对方便的似乎没有太大必要总之都是基于同样硬件架构的。NVIDIA认为目前对于并行计算而言开发语言和API并不是太多而是太少未来DirectX11 Compute Shader同样是基于CUDA架构的。 OPENCL的发布仅仅是一个开始从路线图来看要到2009年中OPENC 1.0正式版才将发布这个时间似乎要比苹果Snow Leopard操作系统发布时间还稍晚些但基于NVIDIA CUDA架构的OPENCL开发的软件在新一代MacOS上有怎样的表现却非常值得期待。 OPENCL的发展蓝图 CUDA C语言的路线图 CUDA C语言路线图则非常细致但在这里NVIDIA没有公布每个版本带来的变化。但纵观半年来CUDA2.0发布后的变化在初期第三方和NVIDIA开发的一些算法已经被CUDA 2.0加入到库中例如FFT高速傅立叶变换FIR滤波等等。对于需要这些数学计算的使用者来说不需要懂得C语言、CUDA编程只要替换库中的相关语句和参数便可以让它的计算提速几十倍。 OPENCL与CUDA的关系大家应该非常清楚了但我们最为好奇的是在中国市场一个看上去仅仅MacOS支持的API会得到如此大关注这有些令人不能理解。虽然苹果的IPOD和IPHONE在中国市场的欢迎程度相当高但MacOS的占有率是相当低即便MacBook的用户也有大半人数和大半时间在使用着Windows操作系统苹果发布OPENCL将在下一代操作系统Snow Leopard中使用竟然有人关心这种边缘新闻 乔布斯为什么会喜欢NVIDIA 经过前文一大串的讲述我们对OPENCL应该有了大概的了解。而我们的身份似乎不经意间变成了苹果和NVIDIA的拥护者大家主动的对OPENCL和CUDA提出各种各样的问题其热情完全不亚于对待3G IPHONE和Macbook Air的发布上市将对待品牌的热情延续到技术品牌与技术本身实在是一件神奇的事情。这不由让我们想起了神奇的乔布斯老乔。 曾经有朋友评价苹果选择NVIDIA非常符合老乔的性格理由很简单老乔喜欢有个性的东西同时也具有对未来敏锐的前瞻性Macbook系列使用NVIDIA GPU产品其实只是一个开始。而我更加同意另一个有趣的观点他们认为老乔这次是被NVIDIA 黄仁勋感动了这也同样很容易理解。当OPENCL在MacOS上得以应用让Macbook因此受益时这才仅仅是NVIDIA CUDA架构与并行计算革命中的一部分CUDA架构的作为不仅仅是OPENCL如此简单CUDA架构目前已经成为并行计算革命的惟一领导者。 从CUDA开始读OpenCL 就像大一学C大二学汇编一样我也写弄了些个月的CUDA然后想想应该开始刨根问底地去学点在CUDA之下层的东西可能会对异构这个编程了解的多。 1 简介 OpenCL全称开发计算语言是并行程序的开发标准使用与任何异构平台——包括多CPU、GPU、CPU与GPU结合等。OpenCL由Khronos Group维护。 OpenCL是一个用于异构平台上编程的开放性行业标准。这个平台可以包括 CPU GPU和其他各类计算设备例如 DSP和Cell/B.E.等等。 OpenCL和CUDA的关系很和谐前者是异构编程规范标准后者是英伟达基于OpenCL之上开发的一个更面向程序员的GPUAPI。所以OpenCL适合于包括英伟达和AMD等的显卡 程序开发。 2 认识OpenCL的框架 2.1 平台模型 [1个host]-[1..N个device] (主机host设备device) [1个device]-[1..N个CU] (计算单元CU) [1个CU]-[1..N个PE] (处理单元PU) host端管理者整个平台的所有计算资源应用程序会从host端向各个 OpenCL设备的处理单元发送计算命令。在一个计算单元内的所有处理单元会执行完全相同的一套指令 流程。指令流可以是 SIMD模式或者SPMD模式。所有由OpenCL编写的应用程序都是从Host启动并结束最终的计算都发生在PE中。 2.2 内存模型 内存介绍 全局内存 (global memory):工作空间内所有的工作节点都可以读写此类内存中的任意元素。OpenCL C提供了缓存global buer的内建函数。 常量内存 (constant memory):工作空间内所有的工作节点可以只读此类内存中的任意元素。 host负责分配和初始化 constant buer在内核执行过程中保持不变。 局部内存 (local memory):从属于一个工作组的内存同一个工作组中所有的工作节点都可以共享使用该类内存。其实现既可以为 OpenCL执行为其分配一块专有内存空间 也有可能直接将其映射到一块global buer上。 私有内存 (private memory):只从属于当前的工作节点。一个工作节点内部的private buer其他节点是完全不可见的。 在这点上基本上和CUDA介绍的内存是一样的。这里的局部内存和CUDA的私有变量差不多一个概念。 内存使用 在内存的使用上有两种方式内存拷贝和内存映射。 拷贝数据是指host通过相应的OpenCL API将数据从host写入到OpenCL设备的内存中或者从 OpenCL设备内存读出数据到 host内存中。 内存映射方法允许用户通过相应 OpenCLAPI将OpenCL的内存对象映射到 host端可见的内存地址空间中。映射之后用户就可以在 host端的映 射地址读写该内存了在读写完成之后用户必须使用对应 API解除这种映射关系。同拷贝内存方式一样映射内存也分block和non-block模式。 2.3 执行模型 OpenCL的执行模型可以分为两部分一部分是在 host上执行的主程序host program另一部分是在 OpenCL设备上执行的内核程序kernelsOpenCL通过主程序来 定义上下文并管理内核程序在OpenCL设备的执行。 执行模式最重要的是分配线程网络这点和CUDA是一回事可以引用。 2.4 编程模型 OpenCL支持按数据并行的编程模型和按任务并行的编程模型。 数据并行模型是指同一系的列指令会作用在内存对象的不同元素上即在不同内存元素上按这个指令序列定义了统一的运算。 在任务并行编程模型是指工作空间内的每个工作节点在执行 kernel程序时相对于其他节点是绝对独立的。在这种模式下对每个工作节点都相当于工作在一个单一的计算单 元内该单元内只有单一工作组该工作组中只有该节点本身在执行。用户可以通过如下方法实现按任务并行: -使用OpenCL设备支持的向量类型数据结构 -同时执行或选择性执行多个kernels -在执行kernels同时交叉性执行一些native kernels程序 OpenCL提供了两个领域的同步 -在同一个工作组中所有的工作节点之间的同步 -同一个上下文中不同的 command queues之间和同一个 command queue的不同commands之间的同步 从CUDA了解openCL是在阅读和理解相关CUDA编程知识后在读《OpenCL中文教程》的一个第一章和第二章的知识汇总去掉了CUDA编程指南中讲解的雷同知识。在了解和 明白OpenCL是怎么一回事儿后我将开始openCL的Hello world了虽然仅仅一个helloword可能没什么意义但是象征性的程序必须写起来。
