在GPU - MATLAB和Simulink上运行CUDA或PTX代码金宝app - 金宝app,下载188bet金宝搏,金宝搏官方网站

运行CUDA或GPU上的PTX代码

概述

本主题介绍如何从CU或PTX（并行线程执行）文件中创建可执行内核，并在MATLAB运行该GPU上的内核^®．核函数在MATLAB中用a表示CUDAKernel对象，可以对MATLAB数组或gpuArray变量进行操作。

以下步骤描述了Cudakernel常规工作流程：

使用编译的PTX代码来创建一个Cudakernel对象，其中包含GPU可执行代码。
设置CUDAKernel对象的属性来控制它在GPU上的执行。
调用函数宏指令在Cudakernel上使用所需的输入，在GPU上运行内核。

遵循这些步骤的MATLAB代码可能看起来像这样:

% 1。创建CUDAKernel对象。k = parallel.gpu.CUDAKernel (“myfun.ptx”，'myfun.cu'，“entryPt1”）;% 2。设置对象的属性。k.GridSize = [8 1];k.ThreadBlockSize = [16 1];％3.使用定义的输入呼叫FeVal。g1 = gpuarray（In1）;％输入GPUArray。g2 = gpuArray (in2);％输入GPUArray。结果=函数宏指令(k, g1, g2);

以下部分提供这些命令和工作流步骤的详细信息。

从CU文件编译一个PTX文件

如果你有一个CU文件想要在GPU上执行，你必须首先编译它来创建一个PTX文件。一种方法是学校网站编译器在NVIDIA^®CUDA^®工具包。例如，如果您的CU文件被调用myfun.cu.，你可以使用shell命令创建一个已编译的PTX文件:

nvcc -ptx myfun.cu

这将生成名为myfun.ptx.．

使用CU文件输入构造CUDAKernel对象

与一个.cu文件和一个.ptx文件，您可以创建CUDAKernel然后，MATLAB中的对象可以使用来评估内核：

k = parallel.gpu.CUDAKernel (“myfun.ptx”，'myfun.cu'）;

请注意

你不能保存或负载Cudakernel对象。

用C原型输入构造CUDAKernel对象

如果您没有与您的PTX文件相对应的CU文件，您可以为您的C内核指定C原型，而不是CU文件。例如:

k = parallel.gpu.CUDAKernel (“myfun.ptx”，'float *， const float *， float'）;

另一个用于C原型输入的用途是源代码使用无法识别的数据类型的重命名时。金宝app（请参阅下面的支持金宝app类型。）假设您的内核包含以下代码。

typedef ArgType浮动;__global__ void add3(ArgType * v1, const ArgType * v2) {int idx = threadadidx .x;[idx] v1 + v2 = [idx];｝

ArgType本身不被识别为支持的数据类型，因此在MATLAB中创建CUD金宝appAKernel对象时，包含它的CU文件不能直接作为输入。控件支持的输入类型金宝appAdd3.内核可以指定为Cudakernel构造函数的C原型输入。例如:

k = parallel.gpu.CUDAKernel (“test.ptx”，'float *， const float *'，'Add3'）;

金宝app支持的数据类型

下表列出金宝app了受支持的C/ c++标准数据类型。

浮点数类型整数类型布尔和字符类型

浮点数类型	整数类型	布尔和字符类型
`双倍的`，`双重` `漂浮`，`float2`	`短的`，`无符号短`，`short2`，`ushort2` `int`，`unsigned int.`，`int2`，`uint2.` `长`，`无符号长`，`朗2`，`ulong2` `很久很久`，`无符号长时间长`，`longlong2`，`ulonglong2` `ptrdiff_t`，`size_t`	`BOOL.` `char`，`无符号的字符`，`char2`，`uchar2`

双倍的，双重

漂浮，float2

短的，无符号短，short2，ushort2

int，unsigned int.，int2，uint2.

长，无符号长，朗2，ulong2

很久很久，无符号长时间长，longlong2，ulonglong2

ptrdiff_t，size_t

BOOL.

char，无符号的字符，char2，uchar2

此外，当您包含时，支持以下整数类型金宝apptmwtypes.h程序中的头文件。

整数类型

整数类型
`INT8_T.`，`int16_T`，`int32_T`，`INT64_T.` `uint8_t.`，`uint16_T`，`uint32_T`，`uint64_T`

INT8_T.，int16_T，int32_T，INT64_T.

uint8_t.，uint16_T，uint32_T，uint64_T

标题文件已作为matlabroot/走读生/ include / tmwtypes.h．你在你的程序中包含这个文件:

#include“tmwtypes.h”

参数的限制

所有的输入都可以是标量或指针，并且可以被标记const．

内核的C声明总是这样的:

__global__ void kernel (input…)

内核必须不返回任何内容，并且只能对其输入参数(标量或指针)进行操作。
内核无法分配任何形式的内存，因此必须在执行内核之前预先分配所有输出。因此，必须在运行内核之前已知所有输出的大小。
原则上，传递到内核的所有指针都是不存在的const可以包含输出数据，因为内核中的许多线程可以修改这些数据。

在将C内核的定义转换为MATLAB时:

C中的所有标量输入(双倍的，漂浮，int，等等)必须是MATLAB中的标量，或者标量(即单元素)gpuArray变量。
所有constC中的指针输入(const双*等）可以是matlab中的标量或矩阵。它们被投射到正确的类型，复制到设备上，并将指向第一元素的指针传递给内核。没有关于原始大小的信息将传递给内核。就好像内核直接收到了结果mxGetData在一个mxArray．
C中的所有非合作指针输入完全将C输入到内核中。但是，由于内核可以更改非合作指针，所以这将被视为来自内核的输出。
来自MATLAB工作区标量和数组的输入将投入请求类型，然后传递给内核。但是，GPUARRAY输入不会自动演绎，因此它们的类型和复杂性必须与预期的类型完全匹配。

这些规则有一些含义。最值得注意的是，来自内核的每个输出必须必须是内核的输入，因为输入允许用户定义输出的大小（从不能在GPU上分配存储器）。

Cudakernel对象属性

当您创建一个没有终止分号的内核对象时，或者当您在命令行中输入对象变量时，MATLAB将显示内核对象属性。例如:

k = parallel.gpu.CUDAKernel ('conv.ptx'，“conv.cu”）

k = parallel.gpu.cudakernel句柄包：parallel.gpu属性：threadblocksize：[1 1] maxthreadsperblock：512 gridsize：[1 1] sharedmemorysize：0ctentpoint：'_z8theentrypf'maxnumlhsguments：1 numrhsarguments：2 ArgumentTypes：{''在单载体'''inout单载体'}

内核对象的属性控制它的某些执行行为。使用点符号来改变那些可以改变的属性。

有关对象属性的描述，请参见CUDAKernel对象引用页面。修改可设置属性的一个典型原因是指定线程的数量，如下所述。

指定入口点

如果您的PTX文件包含多个入口点，您可以在myfun.ptx.你想要内核对象k参考：

k = parallel.gpu.CUDAKernel (“myfun.ptx”，'myfun.cu'，“myKernel1”）;

一个PTX文件可以包含到不同内核的多个入口点。每个入口点都有一个唯一的名称。这些名称通常是被打乱的(如c++中的打乱)。然而，当生成时学校网站PTX名称总是包含CU文件中的原始函数名。例如，如果CU文件将内核函数定义为

__global__ void simplestkernelever（float * x，float val）

则PTX代码包含一个可能被调用的条目_Z18simplestKernelEverPff．

当您有多个入口点时，请在调用时指定特定内核的条目名称CUDAKernel来生成内核。

请注意

的CUDAKernel函数在PTX文件中搜索您的条目名，并匹配出现的任何子字符串。因此，您不应该将任何条目命名为任何其他条目的子字符串。

您可能无法控制原始的条目名称，在这种情况下，您必须知道每个条目的惟一混乱派生。例如，考虑以下函数模板。

template  __global__ void add4(T * v1, const T * v2) {int idx = threadadidx .x;[idx] v1 + v2 = [idx];｝

当模板展开为float和double时，会得到两个入口点，两个入口点都包含子字符串Add4.．

模板__global__ void add4(float *， const float *);模板__global__ void add4(double *， const double *);

PTX有相应的表项:

_Z4add4IfEvPT_PKS0_ _Z4add4IdEvPT_PKS0_

使用入口点Add4.IF.对于浮动版本，和add4id.对于双版本。

k = parallel.gpu.CUDAKernel (“test.ptx”，'double *， const double *'，“add4Id”）;

指定线程数

通过设置其两个对象属性，您可以指定Cudakernel的计算线程数：

GridSize-由三个元素组成的向量，其乘积决定块的数目。
threadblocksize.-一个由三个元素组成的向量，它们的乘积决定了每个块的线程数。(注意，产品不能超过属性的值MaxThreadsPerBlock．）

这两个属性的默认值是(1 1 1)，但是假设您想要使用500个线程并行地对包含500个元素的向量运行与元素相关的操作。一个简单的方法是创建你的CUDAKernel并相应地设置它的属性:

k = parallel.gpu.CUDAKernel(“myfun.ptx”、“myfun.cu”);k.ThreadBlockSize =(500年,1,1);

通常，您可以根据输入的大小设置网格和线程块的大小。有关线程层次结构和多维网格和块的信息，请参阅NVIDIA CUDA C编程指南。

跑一个扶手

使用工作空间变量
使用gpuArray变量
确定输入和输出对应

使用函数宏指令函数来评估GPU上的CUDAKernel。下面的例子展示了如何使用MATLAB工作空间变量和gpuArray变量执行内核。

使用工作空间变量

假设您已经以母语编写了一些内核，并且希望在Matlab中使用它们来在GPU上执行。您有一个内核，这对两个向量进行了卷积;使用两个随机输入向量加载并运行它：

k = parallel.gpu.CUDAKernel ('conv.ptx'，“conv.cu”）;结果= feval（k，rand（100,1），rand（100,1））;

即使输入是MATLAB工作空间数据的常量或变量，输出也是GPUArray.．

使用gpuArray变量

使用起来可能更有效率GPUArray.对象作为运行内核时的输入:

k = parallel.gpu.CUDAKernel ('conv.ptx'，“conv.cu”）;i1 = GPUARRAY（RAND（100,1，'单身的'));i2 = gpuArray (rand(100年1'单身的'));result1 =函数宏指令编写此表达式(k, i1、i2);

因为输出是aGPUArray.，您现在可以使用该输入或输出数据执行其他操作，而无需在MATLAB工作空间和GPU之间进行进一步的传输。当你所有的GPU计算完成时，收集你的最终结果数据到MATLAB工作区:

result2 =函数宏指令(k, i1、i2);r1 =收集(result1)编写此表达式;r2 =收集(result2);

确定输入和输出对应

当调用[out1, out2] = feval(内核，in1, in2, in3)，输入In1.，In2.,In3.对应于CU文件中C函数的每个输入参数。输出OUT1.和out2在执行C内核之后将第一个和第二个非Const指针输入参数的值存储到C函数。

例如，如果一个CU文件中的C内核具有以下签名:

void reallySimple(float * pInOut, float c)

对应的内核对象(k）在Matlab中有以下属性：

MaxNumLHSArguments: 1 NumRHSArguments: 2 ArgumentTypes: {'inout single vector' ' 'in single scalar'}

因此，要使用此代码中的内核对象函数宏指令，你需要提供函数宏指令两个输入参数(除了内核对象)，你可以使用一个输出参数:

y =函数宏指令(k, x1, x2)

输入值X1和X2对应于引出线和c在C函数原型中。输出参数y对应于的值引出线在C内核执行后的C函数原型中。

下面是一个稍微复杂一点的例子，它展示了const和非const指针的组合:

void morecomplicated（const float * pin，float * pinout1，float * pinout2）

MATLAB中相应的内核对象则具有以下属性:

maxnumlhsararguments: 2 numrhsararguments: 3 ArgumentTypes: {'in single vector' 'inout single vector' 'inout single vector'}

您可以使用函数宏指令在这个代码的内核(k)的语法:

(y1, y2) =函数宏指令(k, x1, x2, x3)

三个输入参数X1，X2,X3，对应于传递到C函数的三个参数。输出参数y1.和Y2.，对应于值pInOut1和pInOut2在C内核执行之后。

完整的内核工作流程

添加两个数字
添加两个矢量
示例与CU和PTX文件

添加两个数字

此示例在GPU中添加了两个双打。您应该安装NVIDIA CUDA Toolkit，并为您的设备提供CUDA的驱动程序。

执行此操作的CU代码如下。
```
__global__ void add1(double *pi, double c) {*pi += c;｝
```
该指令__全球的__指示这是内核的入口点。代码使用一个指针将结果发送进来π，它既是输入也是输出。将此代码放入名为test.cu在当前目录中。
在shell命令行中编译CU代码，生成一个名为test.ptx．
```
nvcc -ptx test.cu.
```
在MATLAB中创建内核。目前这个PTX文件只有一个条目，所以您不需要指定它。如果你要放入更多的内核，你会指定Add1.的条目。
```
k = parallel.gpu.CUDAKernel (“test.ptx”，“test.cu”）;
```
用两个数字输入运行内核。默认情况下，内核运行在一个线程上。
```
结果=函数宏指令(k, 2、3)
```
```
结果= 5
```

添加两个矢量

这个例子扩展了前面的例子，将两个向量相加。为简单起见，假设向量中有与元素相同数量的线程，并且只有一个线程块。

CU代码与上一个示例略有不同。两个输入都是指针，一个是常量，因为你没有改变它。每个线程只需在其线程索引处添加元素。线程索引必须计算出这个线程应该添加哪个元素(在CUDA编程中，获取这些特定于线程和块的值是一种非常常见的模式)。
```
__global__ void add2(double * v1, const double * v2) {int idx = threadadidx .x;[idx] v1 + v2 = [idx];｝
```
在文件中保存此代码test.cu．
像以前一样编译学校网站．
```
nvcc -ptx test.cu.
```
如果此代码与第一个示例的代码一起放入同一CU文件中，则需要在此时指定输入点名称以区分其。
```
k = parallel.gpu.CUDAKernel (“test.ptx”，“test.cu”，'Add2'）;
```

在运行内核之前，为要添加的向量正确设置线程数。

n = 128;k.threadblocksize = n;In1 =那些（n，1，'gpuarray'）;in2 = 1 (N, 1'gpuarray'）;结果=函数宏指令(k,三机一体,in2);

示例与CU和PTX文件

有关如何使用CUDA并提供CU和PTX文件供您试验的示例，请参见说明GPU计算的三种方法:Mandelbrot集合．

并行计算工具箱文档

金宝app

尝试MATLAB, Si金宝appmulink和其他产品下载188bet金宝搏

得到审判现在