CUAD&&C++学习笔记

一、头文件

1、mex.h

#include "mex.h"

如果想在Matlab中，以Matlab函数的方式调用C程序，那就要用到MEX文件

2、matrix.h

#include "matrix.h"

这个类数据类型是double，包含了常用的矩阵计算

3、string

里面是一些字符串处理函数，注意这个string与string.h完全是两个文件，之间没有关联，这样定义是因为C中已有一个string.h的库文件，而C++是需要兼容C中的库，所以用另外一种特殊方式导入这个文件：

#include <string>

4、iostream

#include <iostream>

iostream 的意思是输入输出流。#include是标准的C++头文件，任何符合标准的C++开发环境都有这个头文件。
同时，还有一个iostream.h文件，它是input output stream的简写，意思为标准的输入输出流头文件。它包含：
（1）cin>>“要输入的内容”
（2）cout<<“要输出的内容”
这两个输入输出的方法需要#include<iostream.h>来声明头文件。
iostream.h与iostream是不同的。#include<iostream.h>是在旧的标准C++中使用。在新标准中，用#include。

5、omp.h

#include "omp.h"

OpenMP编译必须包含头文件<omp.h>.
OpenMP是用于共享内存并行系统的多处理器程序设计的一套指导性编译处理方案(Compiler Directive) 。

6、cuda_runtime.h

CUDA 目前有两种不同的 API：Runtime API 和 Driver API，两种 API 各有其适用的范围。由于 runtime API 较容易使用，一开始我们会以 runetime API 为主。
cuda_runtime.h头文件一般就是Runtime API中，运行时的API和其参数的定义。（如果使用Driver API，驱动API则头文件使用cuda.h）。

7、stdlib.h

stdlib.h头文件即standard library标准库头文件。里面定义了五种类型、一些宏和通用工具函数。
类型例如：size_t、wchar_t、div_t、ldiv_t和lldiv_t；
宏例如：EXIT_FAILURE、EXIT_SUCCESS、RAND_MAX和MB_CUR_MAX等等；
常用函数如：malloc()、calloc()、realloc()、free()、system()、atoi()、atol()、rand()、srand()、exit()等等。

8、sys/time.h

sys/time.h 是Linux系统的日期头文件。 但sys/time.h 通常会包含include “time.h”。

9、stdio.h

stdio 就是指 “standard input & output"（标准输入输出）,所以，源代码中如用到标准输入输出函数时，就要包含这个头文件！例如c语言中的 printf(“%d”,i); scanf(“%d”,&i);等函数。

10、string.h

在使用到字符数组时需要使用。string .h 头文件定义了一个变量类型、一个宏和各种操作字符数组的函数。

11、time.h

事件相关的头文件

12、math.h

数学计算相关函数的头文件，例如：绝对值、取整和取余、三角函数、反三角函数、双曲三角函数、指数和对数、标准化浮点数、多项式、数学错误计算处理等等。

13、device_launch_parameters.h

要使用threadIdx、blockIdx、blockDim等内置变量时要在头文件里导入device_launch_parameters.h
14、

二、一些声明语句

1、using namespace std

namespace:是指标识符的各种可见范围。是C++语言特别重要的特性，当使用第三方供应商提供的库时，为了避免与其他供应商或者用户定义的名字相冲突（命名空间污染），常常将库的内容放置在自己独立的命名空间中。
std:就是C++标准程序库中定义所有标识符的namespace。
详细解释 参考文章

2、typedef unsigned char byte

 typedef unsigned char byte;

在这个说明以后就可以在程序中用BYTE 代替unsigned char了，只是为了编写代码方便而已！因为typedef是变量类型定义命令，这句话的意思就是将unsigned char类型重新定义为byte，相当于为了简便重新给变量的这种类型命名。
在C/C++中char 表示一个字符变量，占8位，但是可以解释为int型的整数。unsigned char表示 0~255 的整数或者对应的字符。

3、enum

enum是计算机编程语言中的一种数据类型----枚举类型。在实际编程中，有些数据的取值往往是有限的，比如一个星期内只有七天，一年只有十二个月，一个班每周有六门课程等等；并为每个值取一个名字，方便在以后的代码中使用，具体语句格式为：

enum typeName{ valueName1, valueName2, valueName3, ...... };

enum是一个新的关键字，专门用来定义枚举类型，这也是它在C语言中的唯一用途；typeName是枚举类型的名字；valueName1, valueName2, valueName3, …是每个值对应的名字的列表。注意最后的分号；不能少。
应用示例代码：

#include <stdio.h>

int main(){
    enum week{ Mon = 1, Tues, Wed, Thurs, Fri, Sat, Sun } day;
    scanf("%d", &day);
    switch(day){
        case Mon: puts("Monday"); break;
        case Tues: puts("Tuesday"); break;
        case Wed: puts("Wednesday"); break;
        case Thurs: puts("Thursday"); break;
        case Fri: puts("Friday"); break;
        case Sat: puts("Saturday"); break;
        case Sun: puts("Sunday"); break;
        default: puts("Error!");
    }
    return 0;
}

具体用法传送门

4、template

定义函数模板或者类模板。假如我们需要写一个简单的交换函数，但对于C++来说，只能做同数据类型的转换，这时我们就需要针对不同类型之间的交换分别写一个函数，十分麻烦。template就是解决这个问题的，他会告诉系统a是Int型，而b也是int型的。下面是定义方式和调用方式，这种调用swap的方式就是隐式调用方式，可以看到不同类型的值可以统一用这一个函数了，不用单独定义不同类型时的函数。

//声明时这样
template<typename T> 
void Swap(T *a, T *b)
{
    T temp = *a;
    *a = *b;
    *b = temp;
}
//调用时这样
int a=10,b=20;
float c=30,d=40;
swap(&a,&b);
swap(&c,&d);

具体用法传送门

5、Cuda Streams–CUDA流

介绍：
（1）流可以看成是在设备上work的一个队列，host端将work加入队列，然后继续添加。设备在资源free时，开始调度streams里面的work。
（2）CUDA的操作也是在流里面，比如Kernel的启动，内存的拷贝。
（3）在同一个流里面的操作是有序的（FIFO），不可以重叠了。
（4）在不同的流里面的操作是无序的，可以重叠。
用法：

cudaStream_t stream; //声明一个stream
cudaStreamCreate(&stream); //分配stream
cudaStreamDestroy(stream); //取消分配的stream，在stream中的work完成后同步host端。

除非是特别指定了stream，所有的调用都放在默认流里。这个默认流通常指的是“Strewam 0”。
参考网页

5、CUDA程序中函数的前缀

这些函数前缀在官方的文档里被称为函数执行环境标识符Function execution space specifiers，也就是他指明了这段函数是在哪里被调用的。

__global__

这个前缀修饰的函数是核函数，这些函数 在GPU上执行 ，但是需要 在CPU端调用。
注意：
（1）修饰的函数必须采用void返回值，并且需要在调用时制定运行的参数 (也就是<<<>>>里的block数和线程数)；
（2）任何对__global__函数的调用都必须指定该调用的执行配置。执行配置定义将用于在该设备上执行函数的网格和块的维度，以及相关的流。
（3）函数是异步的，这也代表着函数没被执行完就返回了控制权，所以测量核函数的时间需要同步操作才能获得准确的结果。
代码示例：

 naive_copyToTwoVolumes << <gid, blk >> >(函数的输入参数)

其中，尖括号作用为线程配置，gid类型若为dim3，指定网格的维度和大小，gid.x * gid.y * gid.z 等于所启动的块数量；blk的类型若为 dim3，指定各块的维度和大小，Db.x * Db.y * Db.z 等于各块的线程数量；这些参数并不是传递给设备代码的参数，而是告诉运行时如何启动设备代码。传递给设备代码本身的参数是放在圆括号中传递的。

6、threadIdx、blockIdx、blockDim和gridDim

threadIdx  //uint3类型，表示一个线程的索引。
blockIdx  //uint3类型，表示一个线程块的索引，一个线程块中通常有多个线程。
blockDim  //dim3类型，表示线程块的大小。
gridDim  //dim3类型，表示网格的大小，一个网格中通常有多个线程块。

以上涉及到线程、线程块、线程格的知识，下面是它们之间关系的示意图。grid表示线程格，一个线程格内包含许多block线程块，而一个线程块内又包含许多线程thread，其中线程块和线程都可以是一维、二维或三维的。

若线程格和线程块都是三维矩阵。这里假设线程格是一个3×4×5的三维矩阵， 线程块是一个4×5×6的三维矩阵，则有：
（1）gridDim
gridDim.x、gridDim.y、gridDim.z分别表示线程格各个维度的大小：

gridDim.x=3
gridDim.y=4
gridDim.z=5

（2）blockDim
blockDim.x、blockDim.y、blockDim.z分别表示线程块中各个维度的大小：

blockDim.x=4
blockDim.y=5
blockDim.z=6

（3）blockIdx
blockIdx.x、blockIdx.y、blockIdx.z分别表示当前线程块所处的线程格的坐标位置
（4）threadIdx
threadIdx.x、threadIdx.y、threadIdx.z分别表示当前线程所处的线程块的坐标位置
通过 blockIdx.x、blockIdx.y、blockIdx.z、threadIdx.x、threadIdx.y、threadIdx.z就可以完全定位一个线程的坐标位置了。
一般的线程索引方式：

int idz = threadIdx.x + blockIdx.x * blockDim.x;

假如现在我们想得到第6个block中第五个线程，则根据定义式有idz=5+6*8，threadIdx.x线程id为5（0开始），blockIdx.x块id为6，blockDim.x表示块的维度，即一共有8个线程块。

若定义的线程是三维的，则继续定义其他方向有：

int idx = threadIdx.y + blockIdx.y * blockDim.y;
int idy = threadIdx.z + blockIdx.z * blockDim.z;

如果不好理解可以看下图三维grig、block的示意图

                  三维grig、block的示意图

7、make_int2

内置的矢量类型：
char, short, int, long, longlong, float, double：
这些是从基本整数和浮点类型派生的矢量类型。 它们是结构，第一，第二，第三和第四个组件分别可以通过字段x，y，z和w访问。 它们都带有一个构造函数，形式为make_ 。
例如：

int2 make int2(int x, int y);

这条语句将创建一个类型为int2 的向量，值为(x, y).

8、__double2loint、__double2hiint

__double2loint()返回一个double的低32位
__double2hiint()返回一个double的高32位

9、thrust::device_vector

Trust 提供了两个vector容器：host_vector 与 device_vector。按照命名规则，host_vector位于主机端，device_vector位于GPU设备端。Trust的vector容器与STL中的容器类似，是通用的容器，可以存储任何数据类型，可以动态调整大小。
代码示例：

thrust::device_vector<float>& ZXY

其中，thrust::告诉编译器在thrust命名空间中查找函数与类；device_vector表示使用的容器位于GPU设备端；定义的容器类型是float；容器中存放的是ZXY的地址。
还有一个应用的例子：

thrust::device_vector<float> d_xds(xds, xds + DNU);

初始化一段内存，d_xds表示初始化的容器名称，其中的xds表示起始指针，xds + DNU表示结束指针。
另外：
（1）std::cerr与std::cout的不同
cout对应于标准输出流，默认情况下是显示器。这是一个被缓冲的输出，可以被重定向。
cerr对应标准错误流，用于显示错误消息。默认情况下被关联到标准输出流，但它不被缓冲，也就是说错误消息可以直接发送到显示器，而无需等到缓冲区或者新的换行符时，才被显示。一般情况下不被重定向。
参考文章
（2）其中::符号是作用域运算符，假如有下面的代码：

std::cout<<"两个冒号代码的是作用域操作符"<<std::endl;

其中的std就相当于文件夹，count就是里面的文件，<<"两个冒号代码的是作用域操作符"和<<std::endl都是参数，由<<开头表示。其中“”号内的内容就会被输出。

10、const int

const int表示定义一个常量，const int与int不同的点有：
（1）const int类型一旦定义以后就不能修改，int类型是随时可以修改的。
（2）const int 定义的常量不能在其他地方被重新赋值了
例如：const int Max=100; Max++会产生错误;
（3）const int & 是返回这个数值的一个常量的引用，而int 是返回这个数值的一个拷贝。
另外const和int先后顺序可换，即const int=int const。
更多知识点传送门1 传送门2

11、resize

resize()，设置大小（size）;
例子：

const int nsiz_ZXY = (ZN + 1) * (XN + 1) * YN;
ZXY.resize(nsiz_ZXY);

该段代码先定义了常量整数nsiz_ZXY，取nsiz_ZXY的大小对容器ZXY的大小进行设置。
扩展传送门

12、dim3

封装三维数组，例如下面代码行：

dim3 blk(64, 16, 1);

定义了一个名为blk的数组，数组大小为64×16×1。

13、.clear()

代码示例：

thrust::device_vector<float> vol(hvol, hvol + siz);
vol.clear();

其中vol为一个容器，第二行代码clear()表示清楚其中数据，但容器整体结构并未改变。

14、isNormalized

判断是否符合特定的 Unicode 范式。
示例表示判断bool是否符合特定的 Unicode 范式

bool isNormalized

15、#pragma unroll

在默认情形下，编译器对已知次数的小循环进行展开，#pragma unroll 可以用来控制任意一个给定的循环。但 #pragma unroll必须放在被控制的循环的前面，后面可以带展开次数选项。
后跟参数1则编译器不会展开循环。如果没有参数，并且循环次数是一常数时编译器会将循环完全展开，如果不是常数就根本不会展开。
使用示例：

int main()
{
    int a[100];
 
#pragma unroll 4
 
    for(int i=0;i<100;i++)
    {
        a[i]=i;
    }
    return 0;
}

原理解析：
通过使用#pragma unroll命令，编译器在进行编译时，遇到该命令就会对循环进行展开，比如对一些循环次数比较少的循环

for(int i=0;i<4;i++)
    cout<<"hello world"<<endl;

可以展开为：

cout<<"hello world"<<endl;
 
cout<<"hello world"<<endl;
 
cout<<"hello world"<<endl;
 
cout<<"hello world"<<endl;

这样程序的运行效率会更好，当然，现在大多数编译器都会自动这样优化，而通过使用#pragma unroll命令就可以控制编译器的对循环的展开程度

16、#ifndef和#define

作用：防止文件被多次编译，浪费不必要的时间
参考文章

17、cudaMalloc

在GPU内分配内存，这个函数与CPU中的malloc相似，可以先理解malloc的作用，malloc用法为：
int *a = (int )malloc(nsizeof(int))，返回的是一个int型指针，指向大小为n个int型数据的连续内存地址的首地址，可以理解为a是这个数组的首地址。
cudaMalloc也是十分相似的作用，例如cudaMalloc(float(**)&addr,n*sizeof(float))，意思是在GPU内申明一段n个大小的float型数组，addr这个变量中存的就是用户在GPU中声明的float型数组的首地址。

18、cudaMemcpy

用于在主机（Host）和设备（Device）之间往返的传递数据，使用方式：

主机到设备：cudaMemcpy(d_A,h_A,nBytes,cudaMemcpyHostToDevice)
设备到主机：cudaMemcpy(h_A,d_A,nBytes,cudaMemcpyDeviceToHost)

改行代码表示从h_A存储区中复制nBytes个字节到d_A中。
可以根据最后一个参数看出，源指针和目标指针分别是设备指针-主机指针、主机指针-设备指针。
注意：该函数是同步执行函数，在未完成数据的转移操作之前会锁死并一直占有CPU进程的控制权

19、bool success = true

定义一个布尔型变量flag并初始化为真(true)。
代码示例：

bool success = true;
if (succes) printf("balabala");

20、size_t类型

size_t是标准C库中定义的，它是一个基本的与机器相关的无符号整数的C/C + +类型， 它是sizeof操作符返回的结果类型，该类型的大小可选择。

21、uchar4类型

uchar4是在图形交互时使用的一种数据类型

22、计时方式 CLOCKS_PER_SEC

程序示例：

t1 = clock();
t2 = clock();
time = （double）(t2 - t1) / CLOCKS_PER_SEC;

为什么要用时间差除以CLOCKS_PER_SEC呢？
这个CLOCKS_PER_SEC是VC++6.0中time.h下宏定义的常量。其值为1000。它用来表示一秒钟会有多少个时钟计时单元。也就是说clock()将返回时钟周期，CLOCKS_PER_SEC表示一秒钟有多少个时钟周期，两者相除就得到这段时钟周期经历了多少秒钟。

23、inline内联函数

参考文章

24、cudaMallocPitch

主要解决所需分配的矩阵行大小不是128bytes倍数的问题。函数主要作用与cudamalloc相同，不同的是cudaMallocPitch每行分配256bytes。

cudaMallocPitch(void** devPtr, size_t* pitch, size_t widthInBytes, size_t height)

devPtr：开辟矩阵的数据的头指针
pitch：分配存储器的宽度，以字节为单位(cuda的返回值)
pitch是指设备端分配内存时每行的字节数，其计算方式为将每行元素数乘以每个元素的大小，然后向上对齐到内存对齐的大小。例如：数据宽度是33，数据类型是float，则pitch=334（一个float类型数据占4bytes大小）+内存对齐的大小+1。内存对齐的大小也叫做多分配的字节，一般是32/64，由计算机硬件决定。
width：分配矩阵的列数
height：分配矩阵的行数
在设备上分配widthInBytes * height字节的线性内存，并返回分配内存的指针devPtr。
PS：cudaMallocPitch()就是为了解决每行首地址是否是global memory对齐段的问题，如果用cudaMallocPitch()来分配N=33(即列数为33)的矩阵时（需要33*4=132个字节），那么申请的矩阵每一行大小会变成256个Bytes(0-131为我们需要使用的空间，132-255未使用)，而不是cudaMalloc中的132个Bytes，这样分配以后，每行的首地址将会是与globla memory分段地址对齐的（都是128的整数倍），warp在访问的时候就可以对齐了！因此函数将确保在任何给出的行中对应的指针是连续的。
原因

25、cudaMemcpy2D

函数调用方式：

cudaMemcpy2D(void* dst, size_t dpitch, const void* src, size_t spitch,
size_t width, size_t height, enum cudaMemcpyKind kind);

dst: 拷贝到的目标矩阵内存头指针
dpitch: dst指向的2D数组中的内存宽度，以字节为单位，是cuda为了读取方便，对齐过的内存宽度，可能大于一行元素占据的实际内存。实际值大小为dpitchsizeof(数据类型)。
**src ：**源矩阵内存头指针。
spitch: src指向的2D数组中的内存宽度，以字节为单位，spitchsizeof(数据类型)。
width: src指向的2D数组中一行元素占据的实际宽度。以字节为单位，等于
width*sizeof(type)
**height: ** src指向的2D数组的行数。
**kind： **拷贝数据的方向,从src指向的内存区域拷贝数据到dst指向的内存区域。
kind表示拷贝方向：
cudaMemcpyHostToHost；
cudaMemcpyDeviceToHost；
cudaMemcpyHostToDevice；
cudaMemcpyDeviceToDevice。

26、cudaMemset2D

函数调用方法：

cudaError_t cudaMemset2D(void * devPtr, size_t pitch, int value, size_t width, size_t height)	

devPtr - 指向2D设备内存的指针
pitch - 2D设备内存的字节数=pitch*sizeof（type）
value - 为每个指定内存设置的值
width - 设置的矩阵宽度（列数），以字节为单位，所以是Width of matrix set (columns in bytes)*sizeof（type）
height - 设置的矩阵高度（行数）

27、cudaMemcpyToSymbol

将数据从host拷贝到global，拷贝方式的不同是由目的内存申请的方式决定的。
申请的是device内存，cudaMemcpyToSymbol拷贝就是从host拷贝到global memory。
申请的是constant内存，cudaMemcpyToSymbol拷贝就是从host拷贝到constant memory。
调用：

cudaMemcpyToSymbol	(const char * symbol,
const void * src, size_t count, size_t offset = 0,
enum cudaMemcpyKind kind = cudaMemcpyHostToDevice	 
)	

Parameters:
symbol - 设备上拷贝目标
src - 原内存地址
count - 拷贝的数据比特数
offset - 起始符号偏移量（bytes）
kind - 拷贝传输方式
可选： cudaMemcpyHostToDevice / cudaMemcpyDeviceToDevice.

类似的函数：连接

三、一些常用语句

1、__syncthreads()

__syncthreads()是cuda的内建函数，用于块内线程通信。
参考：https://blog.csdn.net/jyl1999xxxx/article/details/68950846

2、srand((unsigned)time(NULL))

如果想在一个程序中生成随机数序列，需要至多在生成随机数之前设置一次随机种子。 即：只需在主程序开始处调用srand((unsigned)time(NULL)); 后面直接用rand就可以了。不要在for等循环放置srand((unsigned)time(NULL));
参考文章

3、cudaThreadSynchronize();

块内通信:通过共享内存进行通信，块内每个线程都能访问共享存储器，不同块的线程不能通信。
__syncthreads()； 当某个线程执行到该函数时，进入等待状态，直到同一线程块（Block）中所有线程都执行到这个函数为止，即一个__syncthreads()相当于一个线程同步点，确保一个Block中所有线程都达到同步，然后线程进入运行状态。
调用 cudaThreadSynchronize()函数,会使cpu处于等待状态,等待所有的线程都执行完毕.但是,cudaThreadSynchronize()函数并不能在kernel中使用。因为CUDA API和host代码是异步的，cudaDeviceSynchronize可以用来停住CPU并等待CUDA中的操作完成。
也就是说cudaThreadSynchronize()是用来同步线程的，而cudaDeviceSynchronize是用来同步整个设备代码的。
参考网页

4、system(“pause”);

在编写的c++程序中，如果是窗口，有时会一闪就消失了，如果不想让其消失，在程序中添加：
system(“pause”);
注意：不要再return 的语句之后加，那样就执行不到了。

5、extern “C”

extern "C"的主要作用就是为了能够正确实现C++代码调用其他C语言代码。因为在C++出现以前，很多代码都是C语言写的，而且很底层的库也是C语言写的，为了更好的支持原来的C代码和已经写好的C语言库，需要在C++中尽可能的支持C，而extern "C"就是其中的一个策略。

6、主函数int main(int argc, char** argv)

基本是固定写法，下面是具体含义，了解即可。
int argc
这个参数用来表示你在命令行下输入命令的时候，一共有多少个参数。比方说你的程序编译后，可执行文件是test.exe，若执行命令D:\tc2>test
这个时候，argc的值是1；
但是，若执行的命令是D:\tc2>test.exe myarg1 myarg2
argc的值是3。也就是一个命令名再加上两个参数，一共三个参数。
char argv[]
这个参数用来取得你所输入的参数
D:\tc2>test，这个时候，argc的值是1，argv[0]的值是 “test”
D:\tc2>test myarg1 myarg2，这个时候，argc的值是3，argc[0]的值是”test”，argc[1]的值是”myarg1”，argc[2]的值是”myarg2”。

7、cudaSetDevice

显卡选择函数，当计算机中有多个GPU时，可以用这个函数选择应用哪个GPU进行并行计算。

cudaSetDevice(0)

8、CUDA_SAFE_CALL

CUDA中一种保护机制，可以用cudaerror代替。

9、cudaDeviceReset

销毁本进程中所有线程在对应GPU设备上的资源分配与状态。

10、纹理内存

（1）纹理对象的创建
一个纹理对象是用cudaCreateTextureObject()产生的。cudaCreateTextureObject()有4个参数，常用的前三个是
cudaTextureObject_t *texObj：需要生产的纹理对象；
cudaResourceDesc *resDesc：资源描述符，用来获取述纹理数据；
cudaTextureDesc *texDesc：纹理描述符，用来描述纹理参数；

//示例代码，产生纹理对象
cudaTextureObject_t texObj1;

（2）纹理内存的寻址方式
cudaAddressModeClamp：超出范围就用边界值代替，示意： AA | ABCDE | EE
cudaAddressModeBorder：超出范围就用零代替，示意： 00 | ABCDE | 00
cudaAddressModeWrap：重叠模式（循环），示意： DE | ABCDE || AB
cudaAddressModeMirror：镜像模式，示意： BA | ABCDE | ED
（3）滤波方式
filtering mode：滤波模式，定义了fetch返回结果的计算方式。有两种模式：cudaFilterModePoint or cudaFilterModeLinear。
cudaFilterModePoint：点模式，返回最接近的一个点，即最近邻插值。
cudaFilterModeLinear：线性模式，即线性插值，对于一维纹理，两点插值；对于二维纹理，四点插值；对于三维纹理，八点插值。线性模式只有在fetch返回浮点类型数据（注意并非指read mode的归一化浮点模式）下才有效
（4） 纹理内存的访问模式
访问模式有 cudaReadModeNormalizedFloat 和 cudaReadModeElementType 两种。 cudaReadModeNormalizedFloat读取 4 字节整数时会除以 0x8fff（有符号整数）或 0xffff（无符号整数），从而把值线性映射到 [-1.0, 1.0] 区间（有符号整数）或 [0, 1] 区间（无符号整数），读取 2 字节整数时也会发生类似变换，除以 0x8f 或 0xff 。
cudaReadModeElementType则不会发生这种转换，直接读取原数据。
（5）cudaResourceDesc资源描述符和cudaTextureDesc纹理描述符

//资源描述符（固定格式）
cudaResourceDesc resDesc;
memset(&resDesc, 0, sizeof(resDesc));//初始化
resDesc.resType = cudaResourceTypeArray;//指定对应设备内存的形式为 CUDA数组
resDesc.res.array.array = cuArray;//CUDA数组 对应的赋值形式

//纹理描述符
cudaTextureDesc texDesc;
memset(&texDesc, 0, sizeof(texDesc));
texDesc.addressMode[0] = cudaAddressModeClamp;
texDesc.addressMode[1] = cudaAddressModeClamp;
texDesc.filterMode = cudaFilterModePoint;//最近邻插值法
texDesc.readMode = cudaReadModeElementType;//若选用cudaFilterModeLinear,则readMode=cudaReadModeNormalizedFloat
texDesc.normalizedCoords = 1;//对坐标进行归一化

接着使用cudaCreateTextureObject()函数创建纹理对象

//创建纹理对象
cudaTextureObject_t tex = 0;
cudaCreateTextureObject(&tex, &resDesc, &texDesc, NULL);

11、CUDA array

(1)cuda array分配
CUDA array是专为纹理获取使用的，是不能通过其它方式进行访问的，其分配函数包括cudaMallocArray和cudaMalloc3DArray。

//以cudaMalloc3DArray为例：
cudaMalloc3DArray //CUDA数组的分配。	
(	
	struct cudaArray ** 	array,
	const struct cudaChannelFormatDesc * 	desc,
	struct cudaExtent 	extent,
	unsigned int 	flags = 0	 
)	

参数:
array - 指向设备内存中已分配数组的指针
desc - 请求通道格式
extent - 请求的分配大小(元素中的宽度字段)
flags - 扩展的标志(目前必须为0)

cudaMalloc3DArray(宽度，高度，深度) 能够分配1D、2D或3D数组，例如：
如果高度和深度范围都为零，则分配一个1D数组。对于1D数组，有效的范围是{(1,8192)，0,0}。同理若
如果只有深度范围为零，则分配2D数组。对于2D数组，有效的范围是{(1,65536)，(1,32768)，0}。
3D为：{(1, 2048), (1, 2048), (1, 2048)}。

(2)纹理通道声明
作用是在设备内存中分配CUDA数组。该函数有一个独立的C和c++ API (c++ API被重载)。
官方使用格式：

cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<float>();

尖括号内为数组数据类型，小括号内为分配数组是所需要的参数，channelDesc为分配的数组名称；
cudaCreateChannelDesc函数的内部结构：

  struct cudaChannelFormatDesc {
          int x, y, z, w;
          enum cudaChannelFormatKind f;
        };
__host__​cudaChannelFormatDesc cudaCreateChannelDesc ( int  x, int  y, int  z, int  w, cudaChannelFormatKind f )

​cudaChannelFormatDesc应该是CUDA频道格式设置的意思，cudaCreateChannelDesc是具体格式设置函数，cudaChannelFormatKind f为设置的数组类型，类型有以下几种：

cudaChannelFormatKindSigned = 0
Signed channel format   （有符号型）
cudaChannelFormatKindUnsigned = 1
Unsigned channel format   （无符号型）
cudaChannelFormatKindFloat = 2
Float channel format   （浮点型）
cudaChannelFormatKindNone = 3
No channel format  （无格式）

其他的四个参数x,y,z,w，对于C API函数，这些是每个通道的比特数。这些可以是颜色通道、空间维度或者任何你想用的东西。从cuda文档中，“返回格式为f的通道描述符，以及每个组件x、y、z和w的比特数。”
示例代码：

cudaChannelFormatDesc channelDesc = cudaCreateChannelDesc<T>();

表示在CUDA内存中分配数组T
（3）cudaMemcpy3DParms
CUDA 3D内存拷贝参数

//用法
cudaMemcpy3DParms copyParams = { 0 };
prjSize.width = Width;//定义3D尺寸-长宽高
prjSize.height = Height;
prjSize.depth = Depth;

copyParams是定义的结构体的名称，需要自定义。
注意：
创建extent时，在旧版本中宽度Width以字节为单位，即必须乘上sizeof(DTYPE)，新版本已经不用乘以sizeof(DTYPE)。
（4）make_cudaPitchedPtr
pitched pointer工具，也是绑定3D纹理内存的一种工具：
因为在当前的CUDA版本中，3D的线性内存是无法直接绑定到texture memory，一维的可以，因此，需要将数据首先放进一个3D的CUDA array，然后将3D CUDA array绑定到texture memory上。

copyParams.srcPtr = make_cudaPitchedPtr(
		(void*)sourceData, prjSize.width*sizeof(type)
		prjSize.width, prjSize.height);

参数说明：
(void*)sourceData–需要传递的数据
prjSize.width*sizeof(type)–维度参数1，注意这个需要乘以数据类型的字节大小
prjSize.width, prjSize.height–维度参数2、3，3D数据这里就是前两个维度

具体原理为，如果访问数组元素*u[x][y][z]，通过pitched pointer访问则是u_p[x+y*pitch+ z*pitch*height ]。 显然，这里pitch=width，因此当创建pitched pointer时我们需要将width和height作为参数传递给函数make_cudaPitchedPtr()。

在这里尤其要注意的是，pitched pointer指向的array与传统的C语言数组的存储方式不同，C语言访问元素u[x][y]*[z]是通过u[y*width*depth+x*depth+z]。因此为了正确读取所需元素，建议逆序建立pitched pointer，即depth和height：

copyParams.srcPtr = make_cudaPitchedPtr((void*)u, array_depth*sizeof(float), array_depth, array_height);

此时相当于数组u[x][y][z]被转置，在CUDA3D array中对应元素为u[z][y][x]。
PS：有些程序，在数据建立的时候就是建立成ZXY形式，后期再经过处理出一个ZYX形式，这里就不用逆序建立了。按照第一个建立方式即可。
参考博文

（5）cuda释放
cudaFreeArray()释放。

12、trust迭代器

重要，单独看。

13、make_float2

make_float2接收一个float类型s，返回一个float2类型，其中x和y的值都设为s。
强制类型转换，这种强制类型转换就是将float类型的数组从第零个元素开始，每两个相邻元素捏成一个float2类的对象，并且是先x后y，生成的float2类型的数组是原长度的一半。

14、dynamic_cast 类型转换操作符

参考文章

15、错误返回函数

cudaGetLastError：
cudaGetLastError函数用于返回最新的一个运行时调用错误。
其返回的值有很多种可能，常见的：
cudaSuccess：成功无异常
cudaErrorSetOnActiveProcess：由异步编程引起的错误
这表明用户在调用非设备管理操作(分配内存和启动内核是非设备管理操作的例子)初始化CUDA运行后，调用了cudaSetDevice()、cudaSetValidDevices()、cudaSetDeviceFlags()等等设置操作。
详细返回值及解析：参考索引文章
cudaGetErrorString：
对于任何CUDA错误，都可以通过函数cudaGetErrorString函数来获取错误的详细信息。

16、CUDA内部计算函式

包括 __expf、__logf、__sinf、__cosf、__powf 等等。这些函式的速度较快，但精确度不如标准的函式。

四、C++函数

1、strcmp

原型：int strcmp(const char *s1, const char *s2);
功能：用来比较两个字符串
参数：s1、s2为两个进行比较的字符串
返回值：若s1、s2字符串相等，则返回零；若s1大于s2，则返回大于零的数；否则，则返回小于零的数。

2、new float

给指针数据开辟新的内存空间

3、vector中.clear()函数

作用：删除 vector 容器中所有的元素，使其变成空的 vector 容器。该函数会改变 vector 的大小（变为 0），但不是改变其容量。其实就是保存了容器，但清空容器原本定义的大小。一般这么做是为了后边继续使用该容器？

vol.clear();

代码中vol是一个vector，clear()的作用是该容器内存储参数的大小改完0，而不是把其中所有元素清0.
具体参考博文：
https://blog.csdn.net/m0_37957160/article/details/106326379
http://c.biancheng.net/view/6846.html

4、memset

作用:在一段内存块中填充某个给定的值，它是对较大的结构体或数组进行清零操作的一种最快方法
函数原型

void *memset(void *str, int c, size_t n)

具体作用原理：复制字符 c（一个无符号字符）到参数 str 所指向的字符串的前 n 个字符。

5、hypotf

返回所传递的参数平方和的平方根
用法：
double hypot(double x, double y);
float hypot(float x, float y);
long double hypot(long double x, long double y);

6、atan2(y, x)

C++中的double类型反正切函数，返回值为弧度，函数功能是求点（x,y）与（0，0）形成的射线与X正半轴的夹角弧度值。一二象限为正，三四象限为负。函数返回值的范围是[-π, π]，对应角度的范围是[-180, 180]度。

与之相似的还有atan2f(y, x)、atan2d(y, x)，分别对应float类型和int类型。参考

7、取绝对值的函数

（1）abs：
用于整数数据类型（如 int、long）。
返回一个整数类型的绝对值。
不需要包含额外的头文件，通常在 或 <stdlib.h> 中定义。
（2)fabs：
用于双精度浮点数数据类型（double）。
返回一个双精度浮点数的绝对值。
需要包含 或 <math.h> 头文件。
（3）fabsf：
用于单精度浮点数数据类型（float）。
返回一个单精度浮点数的绝对值。
需要包含 或 <math.h> 头文件。

8、push_back函数

将一个新的元素加到vector的最后面，位置为当前最后一个元素的下一个元素。

//在容器中添加10
int num = 10;
vector<int> vec;
vec.push_back(num);

9、std::list

Std::list是一种支持在常量时间内插入元素和从容器中任何地方删除元素的容器。不支持快速随机访问。
详细参考