在python上玩CUDA - Numba 与 Pycuda 的比较

python上的CUDA已经广泛应用在TensorFlow，PyTorch等库中，但当我们想用GPU计算资源实现其他的算法时，不得不自己调用CUDA的python接口完成编程，以下是我在python上，利用GPU完成高斯过程计算的经验。

 

【文首劝退】如果是想用CUDA完成较复杂的功能和算法，还是用C++实现吧。。。python的话我感觉很多已定义好的库无法正常调用，虽然numpy的很多属性和功能仍被支持，但如scikit-learn等库无法正常调用，而且坑比较多（错误提示不明显且有关错误和document网络上资料较少），对新手不友好。本人已有一定的C++开发CUDA的经验，但还是费了不少劲在python+CUDA上。

 

实现平台：

window10

python3.7

CUDA10.0（一开始我笔记本安装的是CUDA8.0，因为在尝试pycuda时找不到8.0对应的pycuda版本，后面卸载并安装为10.0。其实后面找到简单的下载pycuda旧版本的方法，8.0还是可以用的）

Pycuda 2018.1.1

一、概述

基于大家都安装好CUDA的前提基础上，在python上使用cuda编程有两种途径：基于Numba 和 基于pycuda（及skcuda）。总体而言，A. Numba编程较容易，且有较详细的document (https://numba.pydata.org/numba-doc/dev/cuda/overview.html)，入门，实现一些简单功能比较容易。但是编程过程中我发现了一些问题。1）不知是不是版本原因，document上的例子没法在我的平台上跑的通（如无法识别python的cmath库，但还是可以识别math库） 2）有些内存管理的api没法正常运行如和local memory相关的（可能只是我菜，没搞明白怎样用）3）可能因为没有显式的分配GPU内存，只有显式地传输，当传入的数组过大时，在memory copy from device to host时有报错的可能性  B. Pycuda编程比Numba复杂，且pycuda 和 skcuda并不是anaconda自带的库，需要额外安装平台对应的版本。但pycuda有gpu相应的库函数，能用gpu完成简单线性代数，基本数学运算，且能结合C代码完成复杂功能，在数据量大的时候加速性能比Numba要好，因此当所需gpu计算编程比较复杂时，建议使用pycuda。pycuda的初步教程可以参考https://zhuanlan.zhihu.com/p/32062796

 

二、Numba的坑

1. 计算过程需要用到的数组需要传入gpu中， 但float,int可以直接传进gpu，而无需api。

第一个红框就是把X1等numpy数组传入gpu，第二个红框指定gpu计算过程的block，thread个数，最后一个红框用于把结果传回来cpu。

2.Numba装饰的代码不能使用numpy的方法或属性

3.Numba装饰的代码不能使用numpy的向量化计算，必须把数组分开成一个个元素这样来计算

@cuda.jit
def gaussian_kernel(x1, x2, m, n, dim, dist_matrix, result, l=1.0, sigma_f=1.0):
    tidx = cuda.threadIdx.x
    bidx = cuda.blockIdx.x

    for i in range(bidx, m, cuda.gridDim.x):
        for j in range(n):
            for k in range(tidx, dim, cuda.blockDim.x):
                dist_matrix[i][j] += (x1[i][k] - x2[j][k]) ** 2
            result[i][j] = sigma_f ** 2 * math.exp(-0.5 * dist_matrix[i][j] / l ** 2)

4. Numba装饰的代码不允许递归调用方法本身。（这对很多算法而言不是太友好）

5. 使用Numba时可能会遇到关于nvvm的报错，这里提供一个解决方法，把下面两个环境变量添加到pycharm的环境变量中（如图）（注意文件路径要根据自己电脑的路径进行修改，且要提前安装好CUDA（上Nvidia官网下载CUDA及显卡对应的驱动版本即可））

NUMBAPRO_NVVM=C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v10.2\nvvm\bin\nvvm64_33_0.dll;

NUMBAPRO_LIBDEVICE=C:\Program Files\NVIDIA GPU Computing Toolkit\CUDA\v10.2\nvvm\libdevice

总的来看，Numba的使用还是比较简单的，打好一个方法，用Numba装饰器修饰它，然后在其他地方传好数组到gpu中就可以调用了。

 

三、Pycuda的坑

1.使用pycuda和skcuda的第一步就是安装这两个库（毕竟不是anaconda自带的）。但是网上搜pycuda的官网或document，有好几个网站都没有提供不同版本的安装包，直到找到这个网址：https://pypi.org/project/pycuda/#history。这样你就可以找版本对应的来安装。其实还有一个简单的方法就是，在pycharm的terminal窗口，用

pip install pycuda == 2018.1.1

也是可以安装的，但要先更新pip到较新版本。

scikit-cuda的安装及document可以参考它的官网：https://scikit-cuda.readthedocs.io/en/latest/install.html。 这里我没遇到什么坑。

库的版本，CUDA版本，显卡驱动版本这几样如果没对应好，就会报“ ... is not supported on this platform" 错误。

2.下面就是把相关的库import到代码中啦，以下几个库基本是必须要用到的，最后一个skcuda库是线性代数的，里面有编写好的GPU矩阵求逆的方法。高斯过程正好要求逆矩阵。

import pycuda.autoinit
import pycuda.gpuarray as gpuarray
import pycuda.driver as dry
from pycuda.compiler import SourceModule
import skcuda.linalg as sklin

若试运行时有很多warning，还可以用

import warnings
warnings.filterwarnings('ignore')

这个来屏蔽那些warning。。。瞬间世界干净了。

3.值得注意的是，pycuda和Numba不能同时用，因为pycuda需要autoinit来初始化gpu context，但这会导致numba认为gpu is non-primary context（就是被人占用了），所以报错。

4.按照网上已有的教程（譬如：https://zhuanlan.zhihu.com/p/32062796）编写好pycuda的代码。同样有以下几步：

a) 先对gpu分配内存

b) 把numpy数组传入gpu

c) 编写kernel方法（就是mod = SourceModule (""" .... """) 这块）

d) 把gpu数组作为参数，调用kernel方法

e) 从device中把结果拿回来

下面就是踩坑填坑时间

5.首先你得确保kernel方法中的c代码语法等正确无误，因为在SourceModule方法中，它以字符串的形式存在，所以debug主要靠经验（当代码量大的时候。。。笑而不语）。基本上，很多报错都是kernel方法的c代码没写好（很多时候的错误报告都会提到

pycuda.driver.CompileError: nvcc preprocessing failed）。有趣的是，错误报告中提及的.cu文件，原来是python自己在一个文件夹中根据c代码生成了一个临时文件，我的电脑的位置是在C:\Users\HP\AppData\Local\Temp里面，大家有兴趣可以看看。（毕竟在VS里面看代码要比字符串要看得舒服

网上有些资料说，这样的processing错误是由于缺少cl.exe文件的系统变量导致，可以把如

C:\Program Files (x86)\VC\bin;

C:\Program Files (x86)\Microsoft Visual Studio 12.0\VC\lib;

的路径添加到系统的PATH变量中（前提是你要安装好VS的某个版本），然后重启pycharm，并把

这个勾给勾上。

6.当kernel方法没有错误后，还要规范好传入参数的数据类型。如下面三条指令

X1.astype(np.float32)
l = np.float32(self.params['l'])
m32 = np.int32(m)

其中X1是numpy数组（可以搭配driver.In(X1), 变成GPU数组，这一点在很多教程中都有提及），self.params['l']是float, m是int (得到的 l 和 m32 才是我们传入kernel方法的变量，因为很多教程都没提float, int 变量怎样直接传入kernel， 我这里列出来以作补充）

 

这样基本上就没什么问题啦~（只是我没有碰到更多的问题而已啦）

想看高斯过程  gpu CUDA简单实现的同学可以访问我的github，Zhixing1020。