Cuda编程-NPP库

Cuda编程先前有过研究，现在记录下Cuda相关的库使用

0.参考文档
1.NPP简介
- 1.1 头文件
- 1.2 库文件
- 1.3 编译时链接关系
- 1.4 NPP函数的命名方式
- 1.5 General Conventions 一般约定
- 1.6 Image Processing Conventions 图像处理约定
2.举例：NPP实现YUV转BGR
- 2.1 简介
- 2.2 利用NPP函数，实现YUV转BGR
- 2.3 NPP相关函数定义
- 2.4 NPP实现Resize操作
- - 2.4.1 参考连接
  - 2.4.2 简介
  - 2.4.3 函数说明
- 2.5 CUDA库之NPP(四):内存开辟和字节对齐
- - 2.5.1 简介
  - 2.5.2 字节对齐
  - 2.5.3 NPP中的阈值函数来探讨字节对齐

0.参考文档

NVIDIA 2D Image and Signal Processing Performance Primitives (NPP)

CUDA库之NPP入门(一)

1.NPP简介

NVIDIA 2D Image and Signal Processing Performance Primitives

NVIDIA NPP是2D图像和信号处理的CUDA加速库。该库中的主要功能集侧重于图像处理，并广泛适用于这些领域的开发人员。NPP库可以在最大限度地提高灵活性，同时保持高性能。可以采用以下两种方式使用NPP库：

作为一个独立的库，可以最小的工作量向应用程序添加GPU加速
可以添加到开发人员的GPU代码中，作为一个有效互操作的协作库

这两种方法都允许开发人员利用NVIDIA gpu的大量计算资源，同时减少开发时间。

1.1 头文件

其中所有的.h文件都安装在CUDA Toolkit目录的include文件下，其中有：
npp.h
nppdefs.h
nppcore.h
nppi.h
npps.h

1.2 库文件

NPP的功能分为3个不同的库组：

NPPC:一个核心库（NPPC），包含来自npp. h头文件的基本功能以及其他两个库使用的通用功能。
NPPI:图像处理库NPPI。nppi. h头文件或名为“nppi_xxx.h”的各种头文件中的任何函数都捆绑到NPPI库中。
NPPS:信号处理库NPPS。npps. h头文件或名为“npps_xxx.h”的各种头文件中的任何函数都捆绑到NPPS库中。

在Windows平台上，NPP存根库位于CUDA Toolkit的库目录中（匹配的DLL位于CUDA Toolkit的二进制目录中* /bin/nppial64_111_<build_no>.dll // Dynamic image-processing library for 64-bit Windows.）：
/lib/nppc.lib
/lib/nppial.lib
/lib/nppicc.lib
/lib/nppidei.lib
/lib/nppif.lib
/lib/nppig.lib
/lib/nppim.lib
/lib/nppist.lib
/lib/nppisu.lib
/lib/nppitc.lib
/lib/npps.lib

在Linux平台上，动态库位于lib目录中，名称包括主要和次要版本号以及内部版本号

* /lib/libnppc.so.11.1.<build_no> // NPP dynamic core library for Linux

1.3 编译时链接关系

注意：静态NPP库依赖于一个名为cuLIBOS（libculibos.a）的公共线程抽象层库，该库现在作为工具包的一部分分发。因此，当静态库被链接时，必须向链接器提供cuLIBOS。为了尽量减少库加载和CUDA运行时启动时间，建议尽可能使用静态库。 为了提高使用动态库时的加载和运行时性能，NPP提供了一整套NPPI子库。仅链接到包含应用程序使用的函数的子库可以显着提高加载时间和运行时启动性能。一些NPPI函数在内部调用其他NPPI和/或NPPS函数，因此您可能需要根据应用程序进行的函数调用链接到一些额外的库。NPPI子库被分成与NPPI头文件拆分方式相对应的部分。子库列表如下：

NPPC 在头文件nppCore.h中，这是NPP核心库，必须链接到应用应用程序中；
NPPIAL 在头文件 nppi_arithmetic_and_logical_operations.h中，定义了一些算术逻辑运算函数；
NPPICC 在头文件nppi_color_conversion.h中，定义了一些颜色转换(如NV12转RGB)和采样（如上采样下采样）功能；
NPPIDEI 在头文件nppi_data_exchange_and_initialization.h中，定义了一些Data交换和初始化的功能；
NPPIF 在头文件nppi_filtering_functions.h中，定义了滤波等计算机视觉相关的功能；
NPPIG 在头文件nppi_geometry_transforms.h中，几何变换函数在定义在里面；
NPPIM 在头文件nppi_morphological_operations.h中，定义了形态学运算函数；
NPPIST在头文件nppi_statistics_functions.h 和 nppi_linear_transforms.h中，定义了统计学与线性变换函数；
NPPISU在头文件nppi_support_functions.h中，定义了内存支持相关的功能；
NPPITC在头文件 nppi_threshold_and_compare_operations.h中，定义了阈值和比较运算函数。

例如，在Linux上，要使用NPP动态库编译一个小的颜色转换应用程序foo，可以使用以下命令

nvcc foo.c -lnppc -lnppicc -o foo # 除了lnppc不可缺少外，只需要链接lnppicc库

如果要针对静态NPP库进行编译，必须使用以下命令：

nvcc foo.c -lnppc_static -lnppicc_static -lculibos -o foo # lculibos必须添加进来

也可以使用本机主机C++编译器。根据主机操作系统的不同，比如我们需要需要一些附加的库，如pthread或dl。可以在Linux上使用以下命令:

g++ foo.c -lnppc_static -lnppicc_static -lculibos -lcudart_static -lpthread -ldl
-I /include -L /lib64 -o foo

NPP是一个无状态的API，从NPP 6.5开始，NPP在函数调用之间记住的唯一状态是当前流ID，即在最近的nppSetStream（）调用中设置的流ID以及关于该流的一些设备特定信息。默认流ID为0。如果应用程序打算对多个流使用NPP，则应用程序有责任使用下面描述的完全无状态的应用程序管理流上下文接口，或者在希望更改流ID时调用nppSetStream（）。任何不使用应用程序管理流上下文的NPP函数调用将使用最近调用nppSetStream（）和nppGetStream（）设置的流，以及其他不包含应用程序管理流上下文参数的“nppGet”类型函数调用也将始终使用该流。

所有NPP函数都应该是线程安全的。

注意：NPP 12.1是NPP的最后一个版本，它将支持不包含NPP流上下文参数的NPP应用编程接口调用。此外，NPP将很快发布一个应用编程接口变体，提供许多应用编程接口调用的折叠组合参数版本。例如，像nppiAdd_8u_C3R_Ctx（pSrc1，nSrc1Step，pSrc2，nSrc2Step，pDst，nDstStep，oSizeROI，nppStreamCtx）这样的调用将成为nppiAdd_Ctx（NPP_8U，NPP_CH_3，pSrc1，nSrc1Step，pSrc2，nSrcStep2，pDst，nDstStep，oSizeROI，nppStreamCtx）。这使得添加对新数据类型和通道数量的支持变得更简单，并显着减少冗余留档。

1.4 NPP函数的命名方式

任意打开一个图像处理中的npp函数，如：

NppStatus nppiYUV420ToBGR_8u_P3C3R(const Npp8u * const pSrc[3], int rSrcStep[3], Npp8u * pDst, int nDstStep, NppiSize oSizeROI);

nppiYUV420ToBGR_8u_P3C3R有着很长的一段后缀，这些后缀除了表示不同的数据类型外，还用来表示原始数据的操作方法

“A”：如果图像是4通道图像，则为“ A”，这表示结果alpha通道不受图元影响。
“Cn”：图像由n个通道压缩像素组成，其中n可以是1、2、3或4。如RGB三个通道， RGBA四个通道；
“Pn”：图像由n个独立的图像平面组成，其中n可以是1、2、3或4。可以理解为 RRRRRGGGGGGBBBBBB这两排列
“C” ：（在通道信息之后）指示图元仅在颜色通道之一（“感兴趣的通道”）上运行。所有其他输出通道不受原语影响。
”I“：表示“in-place”的方式工作。在这种情况下，图像数据指针通常命名为pSrcDst，以指示图像数据同时用作源和目标。
“ M” 表示“mask操作”。这些类型的图元具有附加的“mask图像”作为输入。目标图像中的每个像素对应于mask图像中的像素。仅处理具有相应非零掩码像素的像素。
“ R” 只在矩形的“感兴趣区域”或“ROI”上操作。所有的ROI原语都有一个NppiSize类型的附加输入参数，它指定原语应该处理的矩形区域的宽度和高度。
“Sfs” 表示在输出结果值之前，通过固定的缩放比例和饱和度进行处理。

这些后缀一般是按照字母顺序出现，上面的后缀总是按字母顺序出现。一个4通道原语不影响阿尔法通道的屏蔽操作，如“AC4IMRSfs”。
nppiYUV420ToBGR_8u_P3C3R则表示3通道8位无符号平面YUV420到3通道8位无符号压缩BGR颜色转换。P3表示plane3即包含了三个独立的通道，也就是yyyyuuvv、C3表示三个通道。

1.5 General Conventions 一般约定

1）Memory Management 内存管理
所有NPP函数的设计都遵循与其他NVIDIA CUDA库（如cuFFT和cuBLAS）相同的准则。也就是说，这些API中的所有指针参数都是设备指针。

使用NPP处理数据所涉及的最基本步骤如下：
使用将输入数据从主机传输到设备：>cudaMemCpy(…)
使用一个或多个NPP函数或自定义CUDA内核处理数据
使用将结果数据从设备传输到主机：>cudaMemCpy(…)

2）Scratch Buffer and Host Pointer 抓取缓冲区和主机指针
NPP的一些原语需要额外的设备内存缓冲区（临时缓冲区）来进行计算，例如信号和图像缩减（Sum、Max、Min、MinMax等）。为了让NPP用户最大限度地控制内存分配和性能，用户有责任分配和删除这些临时缓冲区。首先，这有一个好处，即库不会在用户不知情的情况下分配内存。它还允许重复调用同一原语的开发人员只分配一次临时缓冲区，从而提高性能和潜在的设备内存碎片。

3）Function Naming 函数命名
由于NPP是一个C API，因此不允许不同数据类型的函数重载，NPP命名约定解决了区分同一算法或原语函数不同风格但不同数据类型的需要。这种原语不同风格的消歧是通过包含数据类型和其他消歧信息的后缀完成的。
除了风味后缀之外，所有NPP函数都以字母“npp”为前缀。属于NPP图像处理模块的基元在npp前缀中添加字母“i”，即以“nppi”为前缀。类似地，信号处理原语以“npps”为前缀。

一般命名方案为：

npp<module info><PrimitiveName><data-type info>[<additional flavor info>](<parameter list>)

数据类型信息使用与基本NPP数据类型相同的名称。例如，数据类型信息“8u”意味着原语对Npp8u数据进行操作。
如果原语使用与其生成的类型不同的数据，则这两种类型都将按消耗到生成的数据类型的顺序列出。
为每个NPP模块提供了有关“附加风味信息”的详细信息，因为每个问题域使用不同的风味信息后缀。

4)Integer Result Scaling 整数结果缩放
NPP信号处理和成像基元经常对整数数据进行操作。这种整数数据通常是某种物理大小（例如亮度）的定点分数表示。由于这种表示的定点性质，如果将其视为规则整数，许多数值运算（例如加法或乘法）往往会产生超过原始定点范围的结果。
在结果超出原始范围的情况下，这些函数将结果值钳制回有效范围。例如，16位无符号整数的最大正值为32767。4*10000=40000的乘法运算将超过此范围。结果将被钳制为32767。
在这里插入图片描述
5）Rounding Modes 舍入模式
许多NPP函数需要将浮点值转换为整数。舍入模式枚举列出了NPP支持的舍入模式。并非NPP中所有执行舍入作为其功能一部分的原语都允许用户指定使用的舍入模式。相反，它们使用NPP的默认舍入模式，即NPP_RND_FINANCIAL。

6）Rounding Mode Parameter 舍入模式参数
NPP函数的子集执行舍入作为其功能的一部分，允许用户通过舍入模式类型的参数指定使用哪种舍入模式。

1.6 Image Processing Conventions 图像处理约定

1）Function Naming 函数命名
图像处理相关函数使用许多后缀来指示原语的各种不同风格，而不仅仅是不同的数据类型。风味后缀使用以下缩写：

“A”如果图像是4通道图像，则表示结果alpha通道不受原语的影响。
“Cn”图像由n个通道打包像素组成，其中n可以是1、2、3或4。
“pn”图像由n个单独的图像平面组成，其中n可以是1、2、3或4。
“C”（在通道信息之后）表示原语仅在其中一个颜色通道上运行，即“感兴趣的通道”。所有其他输出通道不受原语的影响。
“I”表示原语“就地”工作。在这种情况下，图像数据指针通常命名为pSrcDst，以指示图像数据同时用作源和目标。
“M”表示“掩码操作”。这些类型的基元有一个额外的“掩码图像”作为输入。目标图像中的每个像素对应于掩码图像中的一个像素。只有具有相应非零掩码像素的像素正在被处理。
“R”表示原语仅在矩形region_of_interest或“ROI”上操作。所有ROI原语都采用NppiSize类型的附加输入参数，该参数指定原语应处理的矩形区域的宽度和高度。有关原语如何在ROI上操作的详细信息，请参阅： ref：‘roi_specification’。
“Sfs”表示结果值在写入之前通过固定缩放和饱和度进行处理。

上面的后缀总是按字母顺序出现。例如。4通道原语不影响具有掩码操作的Alpha通道，就位并具有缩放/饱和度和ROI将具有后缀：“AC4IMRSfs”。

2.举例：NPP实现YUV转BGR

CUDA库之NPP(二):NPP实现YUV转BGR

2.1 简介

本文中的例子，仅适合于 512 ∗ 512倍数的图像，因为npp处理数据时，有字节对齐这个说法。
因为只是一个很简单的Demo，所以本文不检测CUDA、NPP函数的返回码，本文利用Opencv读取BGR图像，并转换成YUV的数据格式，然后将数据从主机（Host）拷贝到设备端（Device），调用nppiYUVToBGR_8u_C3R函数，将内存从设备端拷贝到主机端，再利用Opencv将图像保存出来。

2.2 利用NPP函数，实现YUV转BGR

第一步：利用Opencv读取BGR图像，并转换成YUV数据格式

cv::Mat matBrgImg = cv::imread("./data/Fig0638(a)(lenna_RGB).jpg");
int nWidth = matBrgImg.cols; 
int nHeight = matBrgImg.rows;
int nStep = matBrgImg.step; // 每一行的步长，这里 = nWidth * 3
cv::Mat matYuvImg;
cv::cvtColor(matBrgImg, matYuvImg, CV_BGR2YUV);

第二步：将YUV数据从 host拷贝到 dev端

Npp8u *pu8YUV_dev = NULL;
cudaMalloc((void **)& pu8YUV_dev, nWidth * nHeight * 3 * sizeof(Npp8u));
cudaMemcpy(pu8YUV_dev, (Npp8u *)matYuvImg.data, nWidth * nHeight * 3 * sizeof(Npp8u), cudaMemcpyHostToDevice);

这里的 Npp8u在 nppdefs.h里，就是一个 unsigned char
关于更多npp中数据类型中的定义可见本文末

第三步：在Device上创建存放BGR数据的内存
这里推荐使用Npp8u * nppiMalloc_8u_C3(int nWidthPixels, int nHeightPixels, int * pStepBytes);
当然 cudamalloc函数也可以在Device上开辟内存空间。
npp这个函数，可以看到一个参数是 pStepBytes，这个返回每一行占用字节数，由于本文采用的 512 ∗ 512的图像，所以这个值返回为 512 ∗ 3

NppStatus nppRet = NPP_NO_ERROR;
NppiSize nppSize{nWidth, nHeight};
int nLineStep_npp = 0;
Npp8u *pu8BGR_dev = nppiMalloc_8u_C3(nWidth, nHeight, &nLineStep_npp);
printf("nLineStep_npp = %d \n", nLineStep_npp);

Npp8u * nppiMalloc_8u_C3(int nWidthPixels, int nHeightPixels, int * pStepBytes);
返回一个Npp8u 地址
输入有图像的宽、高
int * pStepBytes返回每行占用字节数，由于本文采用的 512 ∗ 512 512512 512∗512 的图像，所以这个值返回为 512 ∗ 3 512 * 3 512∗3
如果输入图像的宽小于512，这个值会补齐为512（这个512可能和显卡设备有关，笔者实验是1080卡）的最小倍数
比如，输入图像的宽为400， pStepBytes = 1536；
比如，输入图像的宽为513， pStepBytes = 2048；
#include “npp.h”
#include “npps_support_functions.h”
int main()
{
int nWidth = 513;
int nHeight = 400;
int nLineStep_npp = 0;
Npp8u *pu8BGR_dev = nppiMalloc_8u_C3(nWidth, nHeight, &nLineStep_npp);
printf(“nLineStep_npp = %d \n”, nLineStep_npp);
printf(“hello world \n”);
return 0;
}

第四步、利用npp中nppiYUVToBGR_8u_C3R函数将yuv数据转换成BGR

nppRet = nppiYUVToBGR_8u_C3R(pu8YUV_dev, nStep, pu8BGR_dev, nStep, nppSize);
 printf("nppRet = %d \n", nppRet);

NppStatus nppiYUVToBGR_8u_C3R(const Npp8u * pSrc, int nSrcStep, Npp8u * pDst, int nDstStep, NppiSize oSizeROI);函数将YUV转换成BGR
*pSrc 源数据地址
*nSrcStep 源数据的Step，即每行占用字节数
*pDst 目的数据
*nDstStep 目的数据的Step.
oSizeROI 感兴趣区域
return 错误码

第五步、将BGR数据从dev端拷贝到Host端，并验证结果

unsigned char *pu8Bgr_host = NULL;
pu8Bgr_host = (unsigned char *)malloc(  nWidth * nHeight * 3);
memset(pu8Bgr_host, 0, nWidth * nHeight * 3);
cudaMemcpy(pu8Bgr_host, pu8BGR_dev, nWidth * nHeight * 3, cudaMemcpyDeviceToHost);

cv::Mat newimage(nHeight, nWidth, CV_8UC3);
memcpy(newimage.data, pu8Bgr_host, nWidth * nHeight * 3);

cv::imwrite("./yuv2BGR.jpg",newimage );

最后，别忘记释放内存空间

if (NULL != pu8BGR_dev)
{
    nppiFree(pu8BGR_dev);
    pu8BGR_dev = NULL;
}

if (NULL != pu8YUV_dev)
{
    cudaFree(pu8YUV_dev);
    pu8YUV_dev = NULL;
}

if (NULL != pu8Bgr_host)
{
    free(pu8Bgr_host);
    pu8Bgr_host = NULL;
 }

2.3 NPP相关函数定义

1、npp中基础数据类型定义，包括int、float等
具体定义可在 nppdefs.h里查下，以下仅摘录关于数据类型定义的一部分。

/** \defgroup npp_basic_types Basic NPP Data Types
 * @{
 */
typedef unsigned char       Npp8u;     /**<  8-bit unsigned chars */
typedef signed char         Npp8s;     /**<  8-bit signed chars */
typedef unsigned short      Npp16u;    /**<  16-bit unsigned integers */
typedef short               Npp16s;    /**<  16-bit signed integers */
typedef unsigned int        Npp32u;    /**<  32-bit unsigned integers */
typedef int                 Npp32s;    /**<  32-bit signed integers */
typedef unsigned long long  Npp64u;    /**<  64-bit unsigned integers */
typedef long long           Npp64s;    /**<  64-bit signed integers */
typedef float               Npp32f;    /**<  32-bit (IEEE) floating-point numbers */
typedef double              Npp64f;    /**<  64-bit floating-point numbers */

2、npp中返回码的定义，主要用来判断函数是否出错以及出错的原因
当返回码为0时，表示成功运行

typedef enum 
{
    /* negative return-codes indicate errors */
    NPP_NOT_SUPPORTED_MODE_ERROR            = -9999,  
    
    NPP_INVALID_HOST_POINTER_ERROR          = -1032,
    NPP_INVALID_DEVICE_POINTER_ERROR        = -1031,
    NPP_LUT_PALETTE_BITSIZE_ERROR           = -1030,
    NPP_ZC_MODE_NOT_SUPPORTED_ERROR         = -1028,      /**<  ZeroCrossing mode not supported  */
    NPP_NOT_SUFFICIENT_COMPUTE_CAPABILITY   = -1027,
    NPP_TEXTURE_BIND_ERROR                  = -1024,
    NPP_WRONG_INTERSECTION_ROI_ERROR        = -1020,
    NPP_HAAR_CLASSIFIER_PIXEL_MATCH_ERROR   = -1006,
    NPP_MEMFREE_ERROR                       = -1005,
    NPP_MEMSET_ERROR                        = -1004,
    NPP_MEMCPY_ERROR                        = -1003,
    NPP_ALIGNMENT_ERROR                     = -1002,
    NPP_CUDA_KERNEL_EXECUTION_ERROR         = -1000,

    NPP_ROUND_MODE_NOT_SUPPORTED_ERROR      = -213,     /**< Unsupported round mode*/
    
    NPP_QUALITY_INDEX_ERROR                 = -210,     /**< Image pixels are constant for quality index */

    NPP_RESIZE_NO_OPERATION_ERROR           = -201,     /**< One of the output image dimensions is less than 1 pixel */

    NPP_OVERFLOW_ERROR                      = -109,     /**< Number overflows the upper or lower limit of the data type */
    NPP_NOT_EVEN_STEP_ERROR                 = -108,     /**< Step value is not pixel multiple */
    NPP_HISTOGRAM_NUMBER_OF_LEVELS_ERROR    = -107,     /**< Number of levels for histogram is less than 2 */
    NPP_LUT_NUMBER_OF_LEVELS_ERROR          = -106,     /**< Number of levels for LUT is less than 2 */

    NPP_CORRUPTED_DATA_ERROR                = -61,      /**< Processed data is corrupted */
    NPP_CHANNEL_ORDER_ERROR                 = -60,      /**< Wrong order of the destination channels */
    NPP_ZERO_MASK_VALUE_ERROR               = -59,      /**< All values of the mask are zero */
    NPP_QUADRANGLE_ERROR                    = -58,      /**< The quadrangle is nonconvex or degenerates into triangle, line or point */
    NPP_RECTANGLE_ERROR                     = -57,      /**< Size of the rectangle region is less than or equal to 1 */
    NPP_COEFFICIENT_ERROR                   = -56,      /**< Unallowable values of the transformation coefficients   */

    NPP_NUMBER_OF_CHANNELS_ERROR            = -53,      /**< Bad or unsupported number of channels */
    NPP_COI_ERROR                           = -52,      /**< Channel of interest is not 1, 2, or 3 */
    NPP_DIVISOR_ERROR                       = -51,      /**< Divisor is equal to zero */

    NPP_CHANNEL_ERROR                       = -47,      /**< Illegal channel index */
    NPP_STRIDE_ERROR                        = -37,      /**< Stride is less than the row length */
    
    NPP_ANCHOR_ERROR                        = -34,      /**< Anchor point is outside mask */
    NPP_MASK_SIZE_ERROR                     = -33,      /**< Lower bound is larger than upper bound */

    NPP_RESIZE_FACTOR_ERROR                 = -23,
    NPP_INTERPOLATION_ERROR                 = -22,
    NPP_MIRROR_FLIP_ERROR                   = -21,
    NPP_MOMENT_00_ZERO_ERROR                = -20,
    NPP_THRESHOLD_NEGATIVE_LEVEL_ERROR      = -19,
    NPP_THRESHOLD_ERROR                     = -18,
    NPP_CONTEXT_MATCH_ERROR                 = -17,
    NPP_FFT_FLAG_ERROR                      = -16,
    NPP_FFT_ORDER_ERROR                     = -15,
    NPP_STEP_ERROR                          = -14,       /**<  Step is less or equal zero */
    NPP_SCALE_RANGE_ERROR                   = -13,
    NPP_DATA_TYPE_ERROR                     = -12,
    NPP_OUT_OFF_RANGE_ERROR                 = -11,
    NPP_DIVIDE_BY_ZERO_ERROR                = -10,
    NPP_MEMORY_ALLOCATION_ERR               = -9,
    NPP_NULL_POINTER_ERROR                  = -8,
    NPP_RANGE_ERROR                         = -7,
    NPP_SIZE_ERROR                          = -6,
    NPP_BAD_ARGUMENT_ERROR                  = -5,
    NPP_NO_MEMORY_ERROR                     = -4,
    NPP_NOT_IMPLEMENTED_ERROR               = -3,
    NPP_ERROR                               = -2,
    NPP_ERROR_RESERVED                      = -1,
    
    /* success */
    NPP_NO_ERROR                            = 0,        /**<  Error free operation */
    NPP_SUCCESS = NPP_NO_ERROR,                         /**<  Successful operation (same as NPP_NO_ERROR) */

    /* positive return-codes indicate warnings */
    NPP_NO_OPERATION_WARNING                = 1,        /**<  Indicates that no operation was performed */
    NPP_DIVIDE_BY_ZERO_WARNING              = 6,        /**<  Divisor is zero however does not terminate the execution */
    NPP_AFFINE_QUAD_INCORRECT_WARNING       = 28,       /**<  Indicates that the quadrangle passed to one of affine warping functions doesn't have necessary properties. First 3 vertices are used, the fourth vertex discarded. */
    NPP_WRONG_INTERSECTION_ROI_WARNING      = 29,       /**<  The given ROI has no interestion with either the source or destination ROI. Thus no operation was performed. */
    NPP_WRONG_INTERSECTION_QUAD_WARNING     = 30,       /**<  The given quadrangle has no intersection with either the source or destination ROI. Thus no operation was performed. */
    NPP_DOUBLE_SIZE_WARNING                 = 35,       /**<  Image size isn't multiple of two. Indicates that in case of 422/411/420 sampling the ROI width/height was modified for proper processing. */
    
    NPP_MISALIGNED_DST_ROI_WARNING          = 10000,    /**<  Speed reduction due to uncoalesced memory accesses warning. */
   
} NppStatus;

2.4 NPP实现Resize操作

2.4.1 参考连接

CUDA库之NPP入门(三):NPP实现Resize操作

2.4.2 简介

本文主要利用npp实现图像中的resize操作，主要步骤如下：
1、利用Opencv读取图像；
2、将图像数据拷贝到设备端；
3、调用nppiResize函数，实现resize操作
4、将nppiResize后的图像数据拷贝到Mat，并保存验证结果

2.4.3 函数说明

NppStatus 
nppiResize_8u_C3R(const Npp8u * pSrc, int nSrcStep, NppiSize oSrcSize, NppiRect oSrcRectROI,Npp8u * pDst, int nDstStep, NppiSize oDstSize, NppiRect oDstRectROI, int eInterpolation);

nSrcStep 指的是步长,即每行数据所占字节数，一般用opencv读取的图像，步长数都是 W ∗ 3 W * 3 W∗3，
也可以调用matSrc.step来获取步长；
**NppiSize**指感兴趣区域操作，这里赋值为图像的大小即可
**eInterpolation** e开头，显然是个枚举，这里指resize中所使用插值类型

npp支持的插值类型都定义在NppiInterpolationMode这个枚举中，可以在nppdefs.h中查看常见的有最近邻、线程插值、三次插值等。

typedef enum 
{
    NPPI_INTER_UNDEFINED         = 0,
    NPPI_INTER_NN                = 1,        /**<  最近邻插值 */
    NPPI_INTER_LINEAR            = 2,        /**<  线性插值 */
    NPPI_INTER_CUBIC             = 4,        /**<  三次插值 */
    NPPI_INTER_CUBIC2P_BSPLINE,              /**<  Two-parameter cubic filter (B=1, C=0) */
    NPPI_INTER_CUBIC2P_CATMULLROM,           /**<  Two-parameter cubic filter (B=0, C=1/2) */
    NPPI_INTER_CUBIC2P_B05C03,               /**<  Two-parameter cubic filter (B=1/2, C=3/10) */
    NPPI_INTER_SUPER             = 8,        /**<  Super sampling. */
    NPPI_INTER_LANCZOS           = 16,       /**<  Lanczos filtering. */
    NPPI_INTER_LANCZOS3_ADVANCED = 17,       /**<  Generic Lanczos filtering with order 3. */
    NPPI_SMOOTH_EDGE             = (1 << 31) /**<  Smooth edge filtering. */
} NppiInterpolationMode;

通常调用opencv的resize函数，即可实现resize操作
如

cv::resize(matSrc, matDst, cv::Size(nRzW, nRzH));

但当出现图像内存解码在GPU上，总不能从GPU将数据拷贝到host端，再调用opencv的resize函数，再从host端拷贝到device端，再执行模型推断，那中间这个拷贝的过程显然是没有必要的。好在NVIDIA已经提供了nppiResize函数用来实现这个功能；

const int nRzH = 450;
const int nRzW = 800;
void npp_resizeData()
{
    cv::Mat matSrc = cv::imread("./data/Fig0638(a)(lenna_RGB).jpg");
    int nH = matSrc.rows;
    int nW = matSrc.cols;
    int nC = matSrc.channels();
    int nStep = matSrc.step;
    printf("nH = %d, nW = %d, nC = %d, nStep = %d\n", nH, nW, nC, nStep);
    // 1. 将图像数据拷贝到设备端
    Npp8u *pu8srcData_dev = NULL;
    cudaMalloc((void **)&pu8srcData_dev, nH * nW * nC * sizeof(Npp8u));
    cudaMemcpy(pu8srcData_dev, matSrc.data, nH * nW * nC * sizeof(Npp8u), cudaMemcpyHostToDevice);

    // 2. 在设备端开辟空间
    Npp8u *pu8dstData_dev = NULL;
    NppiSize npp_srcSize{nW, nH};
    NppiSize npp_dstSize{nRzW, nRzH};
    cudaMalloc((void **)&pu8dstData_dev, nRzH * nRzW * nC * sizeof(Npp8u));
    cudaMemset(pu8dstData_dev, 0, nRzH * nRzW * nC * sizeof(Npp8u));
    // 3.调用nppiresize函数
    nppiResize_8u_C3R( (Npp8u*)pu8srcData_dev, nStep,  npp_srcSize,  NppiRect{0, 0, nW, nH},
                       (Npp8u*)pu8dstData_dev, nRzW * 3, npp_dstSize, NppiRect{0, 0, nRzW, nRzH},
                       NPPI_INTER_LINEAR );

    // 将resize后的图像内存（设备端）拷贝到host端
    cv::Mat newimage(nRzH, nRzW, CV_8UC3);
    cudaMemcpy(newimage.data, pu8dstData_dev, nRzH * nRzW * 3, cudaMemcpyDeviceToHost);
    if (pu8dstData_dev != NULL)
    {
        cudaFree(pu8dstData_dev);
        pu8dstData_dev = NULL;
    }
    if (pu8srcData_dev != NULL)
    {
        cudaFree(pu8srcData_dev);
        pu8srcData_dev = NULL;
    }
    // 保存图像，验证结果
    cv::imwrite("./rzImage_npp.jpg", newimage);
}

2.5 CUDA库之NPP(四):内存开辟和字节对齐

CUDA库之NPP(四):内存开辟和字节对齐

2.5.1 简介

本文主要利用nppiMalloc 来开辟一块内存，并简单探讨npp中的字节对齐问题：
本文只是Demo演示，不考虑返回码的参数检查，也不考虑Free函数来释放内存。

npp的内存开辟函数定义在 nppi_support_functions.h中，

这里以 nppiMalloc_8u_C1 和 nppiMalloc_32fc_C3函数为例：
nppiMalloc_8u_C1：返回一个8bit unsigned char的内存空间，且通道数为1；
nppiMalloc_32f_C3：返回一个32bit float的内存空间，通道数为3；

/**
 * 8-bit unsigned image memory allocator.
 * \param nWidthPixels Image width.
 * \param nHeightPixels Image height.
 * \param pStepBytes \ref line_step.
 * \return Pointer to new image data.
 */
Npp8u  * 
nppiMalloc_8u_C1(int nWidthPixels, int nHeightPixels, int * pStepBytes);

/**
 * 3 channel 32-bit floating point image memory allocator.
 * \param nWidthPixels Image width.
 * \param nHeightPixels Image height.
 * \param pStepBytes \ref line_step.
 * \return Pointer to new image data.
 */
Npp32f * 
nppiMalloc_32f_C3(int nWidthPixels, int nHeightPixels, int * pStepBytes);

参数 nWidthPixels ： 图像宽度，以像素为单位
参数 nHeightPixels ： 图像高度度，以像素为单位
参数 pStepBytes ：输出值，步长，也就是每行所占的字节数，对应opencv的step，bitMap中的stride

2.5.2 字节对齐

通常，Malloc的所开辟的内存，是全局存储器，也就是普通的显存，所有cuda网格上的操作，都能读写全局存储器中的任意位置，而这个读取时存在延迟的，很容易造成性能瓶颈。

所以，在访问显存时，读取和存储就必须字节对齐，如果没有正确对齐，读写将会被编译器拆分为多次操作，降低访存性能。

而我们上一节内容提到的参数 pStepBytes，这个就是字节对齐后，每行所占的字节数。

这个类似Bitmap中的stride，在bitmap图像中，一般是四字节对齐，而在npp中，则和显卡相关，楼主这里是512。

以如下例子：
用npp创建一个 4 ∗ 2 大小的矩阵，返回的step是 512，也就是每行所占的字节数为 512, 这和我们所创建的 2有一些出入。也就是，npp已经自动帮我们在每行补上了510字节的大小了。

总之，无论我们创建的长度所占多少字节，最后用npp来创建内存时，都会补成512的倍数，这就是NPP中的字节对齐。

	const int nRows = 2;
	const int nCols = 4;
	int nLineStep_npp = 0;
	Npp8u* pu8_npp = nppiMalloc_8u_C1(nRows, nCols, &nLineStep_npp);

	printf("Step = %d\n", nLineStep_npp);//512

字节对齐的优点：计算快，这个和CUDA的架构有关，每次读取数据时，能够提高性能；
缺点：拷贝数据时，比较复杂，需要循环开辟空间

2.5.3 NPP中的阈值函数来探讨字节对齐

前面提到，nppMalloc的内存，会在每行自动补上512倍数的内存，那么我们在拷贝数据时，需要考虑缺省的字节长度么？答案时，当然要考虑！

表面上我们只创建了 2 ∗ 4 ∗ s i z e o f ( f l o a t ) 大小的空间，实际上创建了 2 ∗ s t e p 的空间。
补齐的那些空间，我们自然也就不需要用到了，虽然空间有点浪费，但却提高了内存访问的效率。

本例子主要利用NPP的阈值函数，将大于3的数字，都替换成0。
以下是本例子验证的步骤：
Step1：在CPU下创建一个 2 ∗ 4大小的矩阵
Step2：用 nppiMalloc_32f_C1 函数在Cuda上创建内存，并打印字节对齐后的长度
Step3：将CPU创建的数据，拷贝到CUDA中
Step4：调用 nppiThreshold_Val_32f_C1R 阈值处理函数，获取输出结果
Step5：将CUDA上的结果，拷贝到CPU下，打印验证

	const int nH = 2;
	const int nW = 4;

	float data[nH * nW] = { 1.0, 1.0, 2.0, 2.0, 3.0, 4.0, 5.0, 9.0 };

	//float* pSrc_dev = nullptr;
	//cudaMalloc((void**)&pSrc_dev, nH * nW * sizeof(float));
	//cudaMemcpy(pSrc_dev, data, nH * nW * sizeof(float), cudaMemcpyHostToDevice);

	int nLineStep_npp = 0;
	Npp32f* pSrc_dev = nppiMalloc_32f_C1(nH, nW, &nLineStep_npp);
	printf("Step = %d\n", nLineStep_npp);

	for (int i = 0; i < nH; ++i)
	{
		cudaMemcpy((unsigned char*)pSrc_dev + i * nLineStep_npp,
			(unsigned char*)data + i * nW * sizeof(float),  nW * sizeof(float), cudaMemcpyHostToDevice);
	}

	//unsigned char* pDst_dev = nullptr;
	//cudaMalloc((void**)&pDst_dev, nRows * nCols);
	Npp32f* pDst_dev = nppiMalloc_32f_C1(nH, nW, &nLineStep_npp);

	nppiThreshold_Val_32f_C1R(pSrc_dev, nLineStep_npp,
		pDst_dev, nLineStep_npp,
		NppiSize{ nW , nH }, 3.0, 0.0, NPP_CMP_GREATER);
	//nppiThreshold_LTValGTVal_8u_C1R(pSrc_dev, nCols,
	//	pDst_dev, nCols,
	//	NppiSize{ nCols , nRows }, 2u, 0u, 1u, 5u);
	//nppiThreshold_LTValGTVal_32f_C1IR(pSrc_dev, nCols,
	//	NppiSize{ nCols , nRows }, 2u, 0u, 1u, 5u);

	float* pDst_host = (float*)malloc(nH * nW * sizeof(float));
	for (int i = 0; i < nH; ++i)
	{
		cudaMemcpy((unsigned char*)pDst_host + nW * i * sizeof(float), (unsigned char*)pDst_dev + i * nLineStep_npp,  nW * sizeof(float), cudaMemcpyDeviceToHost);
	}
	//cudaMemcpy(pDst_host, pDst_dev, nH * nW * sizeof(float), cudaMemcpyDeviceToHost);
	for (int i = 0; i < 8; ++i)
	{
		printf("%f ", pDst_host[i]);
	}

	// call api to free memory
	std::cout << "\nhello world \n" << std::endl;
	return 0;