理解 CUDA、Numba、Cupy 等的扩展示例

     2023-03-12     107

关键词:

【中文标题】理解 CUDA、Numba、Cupy 等的扩展示例【英文标题】:Extended example to understand CUDA, Numba, Cupy, etc 【发布时间】:2020-12-16 16:39:16 【问题描述】:

大多数在线可用的 Numba、CuPy 等示例都是简单的数组添加,显示了从 cpu 单核/线程到 gpu 的加速。并且命令文档大多缺乏好的例子。这篇文章旨在提供一个更全面的示例。

提供初始代码here。它是经典元胞自动机的简单模型。最初,它甚至不使用 numpy,只是使用普通 python 和 Pyglet 模块进行可视化。

我的目标是将此代码扩展到特定问题(这将非常大),但首先我认为最好已经针对 GPU 使用进行优化。

game_of_life.py 是这样的:

import random as rnd
import pyglet
#import numpy as np
#from numba import vectorize, cuda, jit

class GameOfLife: 
 
    def __init__(self, window_width, window_height, cell_size, percent_fill):
        self.grid_width = int(window_width / cell_size) # cell_size 
        self.grid_height = int(window_height / cell_size) # 
        self.cell_size = cell_size
        self.percent_fill = percent_fill
        self.cells = []
        self.generate_cells()
  
    def generate_cells(self):
        for row in range(0, self.grid_height): 
            self.cells.append([])
            for col in range(0, self.grid_width):
                if rnd.random() < self.percent_fill:
                    self.cells[row].append(1)
                else:
                    self.cells[row].append(0)
                
    def run_rules(self): 
        temp = []
        for row in range(0, self.grid_height):
            temp.append([])
            for col in range(0, self.grid_width):
                cell_sum = sum([self.get_cell_value(row - 1, col),
                                self.get_cell_value(row - 1, col - 1),
                                self.get_cell_value(row,     col - 1),
                                self.get_cell_value(row + 1, col - 1),
                                self.get_cell_value(row + 1, col),
                                self.get_cell_value(row + 1, col + 1),
                                self.get_cell_value(row,     col + 1),
                                self.get_cell_value(row - 1, col + 1)])
                
                if self.cells[row][col] == 0 and cell_sum == 3:
                    temp[row].append(1)
                elif self.cells[row][col] == 1 and (cell_sum == 3 or cell_sum == 2):
                    temp[row].append(1)
                else:                 
                    temp[row].append(0)
        
        self.cells = temp

    def get_cell_value(self, row, col): 
        if row >= 0 and row < self.grid_height and col >= 0 and col < self.grid_width:
           return self.cells[row][col]
        return 0

    def draw(self): 
        for row in range(0, self.grid_height):
            for col in range(0, self.grid_width):
                if self.cells[row][col] == 1:
                    #(0, 0) (0, 20) (20, 0) (20, 20)
                    square_coords = (row * self.cell_size,                  col * self.cell_size,
                                     row * self.cell_size,                  col * self.cell_size + self.cell_size,
                                     row * self.cell_size + self.cell_size, col * self.cell_size,
                                     row * self.cell_size + self.cell_size, col * self.cell_size + self.cell_size)
                    pyglet.graphics.draw_indexed(4, pyglet.gl.GL_TRIANGLES,
                                         [0, 1, 2, 1, 2, 3],
                                         ('v2i', square_coords))

首先,我可以在generate_cells 这个self.cells = np.asarray(self.cells) 的末尾和run_rules 这个self.cells = np.asarray(temp) 的末尾使用numpy 添加,因为之前这样做不会带来加速,如here 所示。(实际上更改为 numpy 并没有带来明显的加速)

例如,关于gpu,我在每个函数之前添加了@jit,并且变得非常慢。 也尝试使用@vectorize(['float32(float32, float32)'], target='cuda'),但这提出了一个问题:如何在只有self 作为输入参数的函数中使用@vectorize

我也试过用numpy代替cupy,比如self.cells = cupy.asarray(self.cells),但也变得很慢。

按照 gpu 使用扩展示例的初步想法,解决问题的正确方法是什么?放置修改/矢量化/并行化/numba/cupy等的正确位置在哪里?最重要的是,为什么?

附加信息:除了提供的代码,这里是 main.py 文件:

import pyglet
from game_of_life import GameOfLife 
 
class Window(pyglet.window.Window):
 
    def __init__(self):
        super().__init__(800,800)
        self.gameOfLife = GameOfLife(self.get_size()[0],
                                     self.get_size()[1],
                                     15,  # the lesser this value, more computation intensive will be
                                     0.5) 

        pyglet.clock.schedule_interval(self.update, 1.0/24.0) # 24 frames per second
 
    def on_draw(self):
        self.clear()
        self.gameOfLife.draw()
        
    def update(self, dt):
        self.gameOfLife.run_rules()
 
if __name__ == '__main__':
    window = Window()
    pyglet.app.run()

【问题讨论】:

我对使用 cuda.jit 装饰器的理解非常有限,但在我看来,这种内核性能不佳的主要原因是在 CPU 和 GPU 之间传输过多数据时。为避免这种情况,必须只传递必要的变量,尤其是在谈论大型数组时。我认为通过使用 self 作为每个函数(将是内核)的参数,您可能会传递不必要的数据。另外,请记住,每个线程都对数组的单个元素进行操作,因此使用 for 迭代数组将不会被并行化。希望这会有所帮助。 @boi ,感谢您指出这一点。我在 3 个月前开始使用 Python,这是我使用的第一种语言。我从来没有使用过,它对我来说是新的,即使我编码了 +10 年。 self_init_ 等对我来说是新的。我会更仔细地查看以正确传递参数。关于for,不知道Python有没有类似parfor的东西,比如Matlab? 实际上是的,numba.prange 可能就是您要找的东西,尽管我认为不可能在 numba.cuda 中并行化循环。这是文档:numba.readthedocs.io/en/stable/user/…。我对这一切也很陌生:)。 @rod_CAE 对此有何更新?我还可以欣赏class 构造的新颖性。 @Sterling 不幸的是没有。该项目在没有并行化的情况下继续进行了一段时间,但由于我的其他项目变得比这个更重要,它目前被搁置了...... 【参考方案1】:

我不太了解您的示例,但我只需要 GPU 计算。痛了几天,大概明白它的用法了,给大家演示一下,希望对大家有所帮助。 另外需要指出的是,在使用“...kernel(cuts,cuts”的时候,我会放两个。因为第一个在传入的时候指定了类型,所以会被核心用作遍历元素,不能被索引读取。所以我用第二个来计算空闲索引数据。

```
binsort_kernel = cp.ElementwiseKernel(
'int32 I,raw T cut,raw T ind,int32 row,int32 col,int32 q','raw T out,raw T bin,raw T num',    
'''
int i_x = i / col;                
int i_y = i % col;                
int b_f = i_x*col;                
int b_l = b_f+col;                
int n_x = i_x * q;                
int inx = i_x%row*col;            
////////////////////////////////////////////////////////////////////////////////////////
int r_x = 0; int adi = 0; int adb = 0;  
////////////////////////////////////////////////////////////////////////////////////////
if (i_y == 0)

for(size_t j=b_f; j<b_l; j++)
    if (cut[j]<q)                
        r_x = inx + j -b_f;       
        adb = n_x + cut[j];       
        adi = bin[adb] + num[adb];
        out[adi] = ind[r_x];      
        num[adb]+= 1;             
    

////////////////////////////////////////////////////////////////////////////////////////
''','binsort')

binsort_kernel(cuts,cuts,ind,row,col,q,iout,bins,bnum)

【讨论】:

这个答案解决了什么问题? OP 的帖子中似乎没有使用 binsort,但看起来您正在使用 CuPy 并展示了如何将 ElementwiseKernel 与 C 代码一起使用?

即使对于巨型矩阵,NUMBA CUDA 也比并行 CPU 慢

...:01:02【问题描述】:网上只有少数几个使用cuda进行numba的示例,我发现它们都比并行CPU方法慢。带有CUDA目标和模板的矢量化甚至更糟,所以我尝试创建一个自定义内核。您随处可见的一篇博文是https://gist.gith 查看详情

可以在用户创建的 numba CUDA 设备函数中调用 numba.cuda.random 设备函数吗?

】可以在用户创建的numbaCUDA设备函数中调用numba.cuda.random设备函数吗?【英文标题】:Canyoucallthenumba.cuda.randomdevicefunctioninuser-creatednumbaCUDAdevicefunctions?【发布时间】:2022-01-0812:19:48【问题描述】:我在numba中有一个cuda内核和几个... 查看详情

Cupy 找不到 CUDA 存储库

】Cupy找不到CUDA存储库【英文标题】:CupyunabletofindCUDArepository【发布时间】:2020-04-2807:35:29【问题描述】:我正在尝试使用cupy,并且我已经安装了CUDA10.2和相应版本的cupy,但是当我尝试运行我的代码时出现此错误:Traceback(mostrece... 查看详情

与 numba.cuda.to_device() 相比,使用 numba.cuda.local.array() 对性能有何影响?

】与numba.cuda.to_device()相比,使用numba.cuda.local.array()对性能有何影响?【英文标题】:Howisperformanceaffectedbyusingnumba.cuda.local.array()comparedwithnumba.cuda.to_device()?【发布时间】:2021-10-2211:00:18【问题描述】:在NumbaDocs:MemoryManagement下,对 查看详情

cuda 与 cupy 和 tensorRT 的流同步问题

】cuda与cupy和tensorRT的流同步问题【英文标题】:CudastreamssynchronizationissuewithcupyandtensorRT【发布时间】:2021-11-1615:37:43【问题描述】:我正在使用TensorRT和cupy。如果我设置了cp.cuda.Stream(non_blocking=True),以下代码也不会等待执行cuda... 查看详情

使用 Numba 进行矩阵乘法时出现 CUDA 内存不足错误

】使用Numba进行矩阵乘法时出现CUDA内存不足错误【英文标题】:CUDAoutofmemoryerrorwhendoingmatrixmultiplicationusingNumba【发布时间】:2021-07-0815:29:44【问题描述】:我需要将一个矩阵与其转置相乘,但我的GPU上的内存不足并出现错误消息n... 查看详情

使用 ctypes 将 cupy 指针传递给 CUDA 内核

】使用ctypes将cupy指针传递给CUDA内核【英文标题】:PassingacupypointertoaCUDAkernelusingctypes【发布时间】:2020-08-1818:12:12【问题描述】:我有一个CUDA内核-template<typenameT,typenameC>__global__voidcuda_ListArray_num(C*tonum,constT*fromstarts,constT*f 查看详情

如何使用 Python 和 Numba 获取 GPU 中的 CUDA 内核数量?

】如何使用Python和Numba获取GPU中的CUDA内核数量?【英文标题】:HowcanIgetthenumberofCUDAcoresinmyGPUusingPythonandNumba?【发布时间】:2020-12-2813:55:22【问题描述】:我想知道如何使用Python、Numba和cudatoolkit获取我的GPU中的CUDA核心总数。【问... 查看详情

如何更新 cupy/CUDA 以使其再次工作并修复我的 conda 环境?

】如何更新cupy/CUDA以使其再次工作并修复我的conda环境?【英文标题】:HowdoIupdatecupy/CUDAsothatitworksagainandfixesmycondaenvironment?【发布时间】:2020-08-2000:51:25【问题描述】:我在Ubuntu18.04(64位)上尝试使用一个函数,该函数使用cupy... 查看详情

如何在 numba CUDA 中对行进行切片?

】如何在numbaCUDA中对行进行切片?【英文标题】:HowtoslicerowsinnumbaCUDA?【发布时间】:2021-08-1504:13:09【问题描述】:我是Numba的初学者。我很难在GPU中重新排列数组的行。例如,在NumbaCPU中,这可以通过fromnumbaimportnjitimportnumpyasnp@n... 查看详情

我可以从c ++调用用cupy代码编写的cuda代码吗?

】我可以从c++调用用cupy代码编写的cuda代码吗?【英文标题】:CanIcallcudacodewrittenincupycodefromc++?【发布时间】:2019-08-1603:21:07【问题描述】:如果我有一个简单的采样3D点的形式s=[[x1,y1,z1],[x2,y2,z2],[x3,y3,z3],[x4,y4,z4],.....]在cupy中实现... 查看详情

为什么这个numba.cuda查找表实现失败?(代码片段)

...卡造成问题。但是下面的代码失败并出现未知错误:fromnumbaimportcuda,vectorizeimportnumpyasnptmp=np.random.uniform(0,100,1000000).astype(np.int16)tmp_device=cuda.to_device(tmp)lut 查看详情

Numba 无法使用完整的 GPU

】Numba无法使用完整的GPU【英文标题】:NumbafailingtousethefullGPU【发布时间】:2021-12-1922:23:06【问题描述】:我最近开始使用Numba作为大学作业的一部分,目的是比较Numba和Cuda-C之间特定的GPU并行代码的性能。我已经用Cuda-C编写了代... 查看详情

Cupy 错误 - 超出磁盘配额 [包括最小示例]

】Cupy错误-超出磁盘配额[包括最小示例]【英文标题】:Cupyerror-diskquotaexceeded[Minimalexampleincluded]【发布时间】:2019-08-0107:36:16【问题描述】:在cupy中执行以下最小示例。importcupy,cupyxx=cupy.array([1.,2.,3.])y=cupy.arange(10)print(cupyx.get_runtim... 查看详情

Numba Cuda 计算似乎比顺序运行慢。我犯了明显的错误吗?

】NumbaCuda计算似乎比顺序运行慢。我犯了明显的错误吗?【英文标题】:NumbaCudacomputationseemstobeslowerthansequentialrun.DidIdoobviousmistakes?【发布时间】:2021-11-2905:44:49【问题描述】:有几个线程涵盖了类似的主题,但不幸的是,这些似... 查看详情

什么是 numba.cuda.local.array() 的有效替代方案,它们不像通过 to_device() 传递许多参数那么麻烦?

】什么是numba.cuda.local.array()的有效替代方案,它们不像通过to_device()传递许多参数那么麻烦?【英文标题】:Whatareefficientalternativestonumba.cuda.local.array()thataren\'tascumbersomeaspassingmanyargumentsviato_device()?【发布时间】:2021-10-2211:44:20【... 查看详情

CuDNN 在二进制安装的 cupy 中不可用

...间】:2019-04-2405:03:31【问题描述】:我使用二进制包cupy-cuda92安装了Cupy,但没有加载CuDNN。$pipinstallcupy-cuda92==5.4.0chainer==5.4.0$python-c\'importchainer;chainer.print_ 查看详情

pythongpu加速数据科学|计算距离矩阵在用cupy时快了约100倍

...da.org/rapidsai/pylibraft也可以用pipinstall安装pylibraft-cu11和cupy-cuda11x(注意:我本地CUDA版本为11.3,因此选择pylibraft-cu11与cupy-cuda11x&# 查看详情