DAY87:阅读Interoperability between Runtime and Driver APIs

I.4. Interoperability between Runtime and Driver APIs

An application can mix runtime API code with driver API code.

If a context is created and made current via the driver API, subsequent runtime calls will pick up this context instead of creating a new one.

If the runtime is initialized (implicitly as mentioned in CUDA C Runtime), cuCtxGetCurrent() can be used to retrieve the context created during initialization. This context can be used by subsequent driver API calls.

Device memory can be allocated and freed using either API. CUdeviceptr can be cast to regular pointers and vice-versa:

In particular, this means that applications written using the driver API can invoke libraries written using the runtime API (such as cuFFT, cuBLAS, ...).

All functions from the device and version management sections of the reference manual can be used interchangeably.

本文备注/经验分享:

今天的章节是Driver API简介的最后一篇,如同我们之前所说的,你需要学会了以前的一些Runtime API中没有的概念:CUDA Context、Module,以及kernel在Driver API下的启动。实际上,除了这三者外,其他的东西基本上只有函数名字的不同,而用法基本一样的(例如cu开头和cuda开头)。 则今天的章节则是Driver API和Runtime API互操作。 在首天的Context章节里面,你已经知道CUDA Context能否实现卡上分配的资源的隔离,同时你也知道runtime api会自动隐式初始化,自动使用一个context的。 那么问题就来了,我能否在同时使用了Driver API和Runtime API的应用里面,能否在Driver API和Runtime API的各自的context里面(例如前者是手工创建的,后者是自动的),将资源互相共享?从而使得Driver API能使用Runtime API里面的东西,或者相反?例如本章节给出了一个问题,我能否使用cublas(基于runtime api),在一个driver api应用里面? 实际上是可以的。 存在多种办法来完成这个共享,一个是最直接想到的笨办法,内存是通用的,于是产生了: Driver API的Context <---> 系统内存 <----> Runtime API的Context 例如你可以将一个Runtime API的Context中有效的buffer中的数据传到内存,在从内存传到另外一个Driver API的Context的buffer中。这种办法,的确可以工作,但是很低效。因为你引入了两次无辜的传输。 但这个是一个直接的思路,例如看我们论坛的: http://bbs.gpuworld.cn/thread-59078-1-2.html http://bbs.gpuworld.cn/thread-59076-1-2.html 类似这两个帖子中的问题。都是很直接的想法。但是实际上,如果这些用户阅读了今天的章节,楼主们会有更好的做法: (1)Driver API和Runtime API可以共享同一个Context: 这主要又分为Driver API先建立Context,并设定成当前线程的当前Context(第一个Driver API中建立该Context的Host线程会自动设定,其他线程需要手工),然后任何后续的Runtime API将取消自行的隐形建立过程,而是就地使用该Context。 和分成:Runtime API先初始化,然后Driver API获取Runtime API所创建的默认Context,并使用它(cuCtxGetCurrent,在本Host线程的之前调用任何常规Runtime API函数后)。 通过这两种方式中的任何一种,将使得该应用的Host进程中,能够让Runtime API和Driver API共享CUDA Context的。这样就规避了数据不能共用的问题 (主要是分配的显存)。 那么除了缓冲区(显存)外,其他的Driver API特有的Module和kernel指针对象,能否在Runtime API中共享,这个倒是不能的。因为Runtime API将全自动管理它们。 但好在很少有需要将同一个kernel用两种不同的方式分别从Runtime和Driver API中启动的情况存在,所以这一般不是个问题。而较大的缓冲区的共享确实额外重要的。所以能否使用同一个context,共享缓冲区中的数据,就一般情况下足够了。 这是第一点。 除了让Runtime API和Driver API共享一个Context,能否互相利用数据外,本章节的互操作还有另外的一个重要的用户。 重要的用途: 随着CUDA的应用越来越广泛,很多第三方的库都在使用CUDA,特别是Runtime API。而这些第三方的库在开发的时候,可能没有注意到和调用者,或者其他的类似的利用了CUDA的库之间的和平共处的问题,例如刚才发的两个帖子中所说的: 很多第三方的库,同时在使用的时候,它们习惯性的假如cudaDeviceReset()(注意你没看错)。从而导致另外一个库提示“Driver正在卸载”,或者“未知错误”之类的奇葩现象,这个就需要你自行处理了。使用第三方的库,那么就要付出使用的代价。 此时通过利用Driver API的这种Context互操作的方式,特别是在我们拿不到这些第三方库的代码,无法对它们进行修改的情况下,可以进行安全的调用: 例如Driver API中可以创建一个空白Context,然后用这个空白Context作为当前Host线程的当前Context,再调用这些第三方的库,这样就可以让它们随意的折腾了。比较安全。 很多类似的用法都可以想到。 实际上,这就是Driver API所带来的基本福利,你拥有更强和更细微的控制力。如同刚才说过的,主要的3大点Driver API引入的不同(Context/Module/Kernel启动),后两个你无法直接在Runtime API里共享,可能很多人也用不到;但是前者的Context的控制能力,所有的CUDA Runtime API用户应当稍微学会一点,还是很有用的。所以这几天的Driver API章节,你至少应当将第一天的,和第二天的内容看完。 例如刚才论坛上的两个问题,再例如很多人问的,为何我在CUDA (Runtime API)的应用中,首次调用某些cuda开头的函数(例如cudaMalloc), 总会卡上一段时间,为何后续的同样的函数调用,就快了很多?这个如果你阅读了这些天的章节,你至少应该知道,这里面存在runtime ap在自动进行初始化context,自动载入module之类的过程,这些过程中在手册调用某些cuda开头的函数,或者cudaDeviceReset()后,都会自动触发,这样你就不会迷茫的四处怀疑了。 实际上,不仅仅是今天章节的内容,用户还应当看一下CUDA Runtime的隐形主Context的概念(网上就有),还应当看一下cuDevicePrimaryCtxt*()开头的driver api函数,获取更详细的信息。 这样到今天,Driver API的简介就告一段落。如同刚才所说,你至少应当看一下前两天的内容,很有用途的。 虽然说Driver API用起来比较麻烦,但如同我们第一天说的,它带来的更大的灵活性,更大的平台和语言适应能力,更好的安全控制(例如今天的例子),都是严肃的CUDA使用者应当考虑的问题。 再次举个例子说,某联机的应用,服务器段发布了一次动态更新(GPU上的kernel代码),如果你使用传统的自动更新方式 + CUDA Runtime API,你只能保存更新到exe,然后下次启动的时候应用,但是如果你使用了Driver API,随时可以从网络传来新的Module,然后你的应用完成在内存中,不停机的,旧Module卸载,新Module载入,外加新的kernel启动的过程,一切都是热更新的,没有停机过程,你看这就是另外一个灵活性的例子。

有不明白的地方,请在本文后留言

本文分享自微信公众号 - 吉浦迅科技(gpusolution)

原文出处及转载信息见文内详细说明,如有侵权,请联系 yunjia_community@tencent.com 删除。

原始发表时间:2018-11-12

本文参与腾讯云自媒体分享计划,欢迎正在阅读的你也加入,一起分享。

发表于

我来说两句

0 条评论
登录 后参与评论

相关文章

来自专栏大数据文摘

迷人又诡异的辛普森悖论:同一个数据集是如何证明两个完全相反的观点的?

在辛普森悖论中,餐馆可以同时比竞争对手更好或更差,锻炼可以降低和增加疾病的风险,同样的数据集能够用于证明两个完全相反的论点。

15930
来自专栏数据结构笔记

python基础类型(一):字符串和列表

注意到最后三个的单双引号是嵌套使用的,但是最后一个的使用方法是错误的,因为当我们混合使用两种引号时必须有一种用来划分字符串的边界,即在两边的引号不能出现在字符串...

14020
来自专栏Python专栏

200行代码,一行行教你自制微信机器人

1) 用一个windows客户端工具运营公众号,真的很局限。虽然工具的功能很强大,能自动添加好友,自动拉好友入群,关键字回复等等,但是有一个绕不开的点,它是一款...

62520
来自专栏smy

一张图解释负载均衡

首先当大量用户访问时候,先请求到nignx服务器,因为nignx对于高并发支持较好,所以由nignx服务器将访问需求分配给不同的apache服务器,apache...

21330
来自专栏编程坑太多

『高级篇』docker之Mesos集群架构图(23)

12140
来自专栏机器之心

Diss所有深度生成模型,DeepMind说它们真的不知道到底不知道什么

深度学习在应用层面获得了巨大成功,这些实际应用一般都希望利用判别模型构建条件分布 p(y|x),其中 y 是标签、x 是特征。但这些判别模型无法处理从其他分布中...

11710
来自专栏苦逼的码农

一些常用的算法技巧总结

数组的下标是一个隐含的很有用的数组,特别是在统计一些数字,或者判断一些整型数是否出现过的时候。例如,给你一串字母,让你判断这些字母出现的次数时,我们就可以把这些...

22030
来自专栏chenssy

多线程:为什么在while循环中加入System.out.println,线程可以停止

这个我们都知道,由于 stopReqested 的更新值在主内存中,而线程栈中的值不是最新的,所以会一直循环,线程并不能停止。加上 Volatile 关键字后,...

21540
来自专栏GreenLeaves

TFS2018环境搭建一硬件要求

TFS可以安装在Windows Server和Windows PC操作系统中,但是TFS2018和2018只支持64位操作系统中,早期的版本没有操作系统的位数限...

41130
来自专栏我是攻城师

理解BitMap算法的原理

位图:一种常用的数据结构,代表了有限域中的稠集(dense set),每一个元素至少出现一次,没有其他的数据和元素相关联。在索引,数据压缩,海量数据处理等方面有...

25430

扫码关注云+社区

领取腾讯云代金券

年度创作总结 领取年终奖励