1.4. FP4转换与数据移动

要使用这些函数，请在程序中包含头文件 cuda_fp4.h。

枚举

__nv_fp4_interpretation_t: 列举将4位值作为fp4类型时的可能解释。

Functions

__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_bfloat16raw2_to_fp4x2(const __nv_bfloat162_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 将输入向量中两个nv_bfloat16精度的数字（打包在__nv_bfloat162_raw x中）转换为两个fp4类型值的向量，使用指定的舍入模式并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4_storage_t __nv_cvt_bfloat16raw_to_fp4(const __nv_bfloat16_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 将输入的nv_bfloat16精度x转换为指定类型的fp4格式，使用指定的舍入模式并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_double2_to_fp4x2(const double2 x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 将输入向量中两个打包在double2 x中的双精度数，使用指定的舍入模式转换为所请求类型的两个fp4值向量，并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4_storage_t __nv_cvt_double_to_fp4(const double x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 将输入的double精度x转换为指定舍入模式的fp4类型，并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_float2_to_fp4x2(const float2 x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 将输入向量中的两个single精度数（打包在float2 x中）转换为两个fp4类型值的向量，使用指定的舍入模式并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4_storage_t __nv_cvt_float_to_fp4(const float x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 使用指定的舍入模式将输入的single精度x转换为请求类型的fp4，并对超出范围的值进行饱和处理。
__host__ __device__ __half_raw __nv_cvt_fp4_to_halfraw(const __nv_fp4_storage_t x, const __nv_fp4_interpretation_t fp4_interpretation): 将指定类型的输入 fp4 x 转换为 half 精度。
__host__ __device__ __half2_raw __nv_cvt_fp4x2_to_halfraw2(const __nv_fp4x2_storage_t x, const __nv_fp4_interpretation_t fp4_interpretation): 将指定类型的两个fp4值的输入向量转换为打包在__half2_raw结构中的两个half精度值的向量。
__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_halfraw2_to_fp4x2(const __half2_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 将输入向量中两个打包在__half2_raw x的半精度数，使用指定的舍入模式转换为两个fp4类型值的向量，并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4_storage_t __nv_cvt_halfraw_to_fp4(const __half_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding): 使用指定的舍入模式将输入的half精度x转换为请求类型的fp4，并对超出范围的值进行饱和处理。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(): 默认构造函数。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const unsigned long int val): 从unsigned long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const double f): 从double数据类型构造，对于超出范围的值依赖于__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const long int val): 从long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const float f): 从float数据类型构造，对于超出范围的值依赖__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const int val): 从int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const unsigned short int val): 从unsigned short int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const long long int val): 从long long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const short int val): 从short int数据类型的构造函数。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const __nv_bfloat16 f): 从__nv_bfloat16数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const unsigned int val): 从unsigned int数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const unsigned long long int val): 从unsigned long long int数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4_e2m1::__nv_fp4_e2m1(const __half f): 从__half数据类型构造，依赖__NV_SATFINITE处理超出范围值的行为，并使用cudaRoundNearest舍入模式。
__host__ __device__ __nv_fp4x2_e2m1::__nv_fp4x2_e2m1(const double2 f): 从double2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x2_e2m1::__nv_fp4x2_e2m1(const __nv_bfloat162 f): 从__nv_bfloat162数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x2_e2m1::__nv_fp4x2_e2m1(const __half2 f): 从__half2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x2_e2m1::__nv_fp4x2_e2m1(const float2 f): 从float2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x2_e2m1::__nv_fp4x2_e2m1(): 默认构造函数。
__host__ __device__ __nv_fp4x4_e2m1::__nv_fp4x4_e2m1(): 默认构造函数。
__host__ __device__ __nv_fp4x4_e2m1::__nv_fp4x4_e2m1(const __nv_bfloat162 flo, const __nv_bfloat162 fhi): 从一对__nv_bfloat162数据类型值构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x4_e2m1::__nv_fp4x4_e2m1(const double4 f): 从double4向量数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x4_e2m1::__nv_fp4x4_e2m1(const float4 f): 从float4向量数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。
__host__ __device__ __nv_fp4x4_e2m1::__nv_fp4x4_e2m1(const __half2 flo, const __half2 fhi): 从一对__half2数据类型值构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

类型定义

__nv_fp4_storage_t: 用于存储fp4浮点数的8位unsignedinteger类型抽象。
__nv_fp4x2_storage_t: 用于存储fp4浮点数对的8位unsignedinteger类型抽象。
__nv_fp4x4_storage_t: 用于存储fp4浮点数四元组的16位unsignedinteger类型抽象。

1.4.1. 枚举

enum __nv_fp4_interpretation_t

列举将4位值作为fp4类型引用时可能的解释方式。

取值：

enumerator __NV_E2M1: 代表fp4类型的e2m1数值。

1.4.2. 功能

__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_bfloat16raw2_to_fp4x2(const __nv_bfloat162_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

将输入向量中两个nv_bfloat16精度的数字（打包在__nv_bfloat162_raw x中）转换为两个fp4类型的值向量，使用指定的舍入模式并饱和处理超出范围的值。

将输入向量x转换为由fp4_interpretation参数指定类型的两个fp4值向量，使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4x2_storage_t 值保存转换的结果。

__host__ __device__ __nv_fp4_storage_t __nv_cvt_bfloat16raw_to_fp4(const __nv_bfloat16_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

将输入的nv_bfloat16精度x转换为指定类型的fp4格式，使用指定的舍入模式并对超出范围的值进行饱和处理。

将输入x转换为由fp4_interpretation参数指定类型的fp4格式，使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4_storage_t 值保存转换的结果。

__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_double2_to_fp4x2(const double2 x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

将输入向量中两个double精度的数字（打包在double2 x中）转换为两个fp4类型值的向量，使用指定的舍入模式并对超出范围的值进行饱和处理。

将输入向量x转换为由fp4_interpretation参数指定类型的两个fp4值向量，使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp4_storage_t __nv_cvt_double_to_fp4(const double x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

将输入的double精度x转换为指定舍入模式的fp4类型，并对超出范围的值进行饱和处理。

将输入x转换为由fp4_interpretation参数指定类型的fp4格式，使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4_storage_t 值保存转换结果。

__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_float2_to_fp4x2(const float2 x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

将输入向量中的两个single精度数字（打包在float2 x中）转换为两个fp4类型值的向量，使用指定的舍入模式并对超出范围的值进行饱和处理。

将输入向量x转换为由fp4_interpretation参数指定类型的两个fp4值组成的向量，使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4x2_storage_t 值保存转换的结果。

__host__ __device__ __nv_fp4_storage_t __nv_cvt_float_to_fp4(const float x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

使用指定的舍入模式将输入的single精度x转换为请求类型的fp4，并对超出范围的值进行饱和处理。

将输入x转换为由fp4_interpretation参数指定类型的fp4格式，并使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4_storage_t 值保存转换的结果。

__host__ __device__ __half_raw __nv_cvt_fp4_to_halfraw(const __nv_fp4_storage_t x, const __nv_fp4_interpretation_t fp4_interpretation)

将指定类型的输入 fp4 x 转换为 half 精度。

将输入x从fp4类型（由fp4_interpretation参数指定）转换为half精度。

Returns

__half_raw 值保存转换结果。

__host__ __device__ __half2_raw __nv_cvt_fp4x2_to_halfraw2(const __nv_fp4x2_storage_t x, const __nv_fp4_interpretation_t fp4_interpretation)

将指定类型的两个fp4值组成的输入向量转换为打包在__half2_raw结构中的两个half精度值组成的向量。

将输入向量 x 从由 fp4_interpretation 参数指定的 fp4 类型转换为两个 half 精度值的向量，并以 __half2_raw 结构返回。

Returns

__half2_raw 值保存转换结果。

__host__ __device__ __nv_fp4x2_storage_t __nv_cvt_halfraw2_to_fp4x2(const __half2_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

将输入向量中两个打包在__half2_raw x的半精度数转换为两个指定类型的fp4值向量，使用指定的舍入模式并饱和处理超出范围的值。

将输入向量x转换为由fp4_interpretation参数指定类型的两个fp4值向量，使用rounding参数指定的舍入模式。超出范围的大数值会饱和至相同符号的MAXNORM值。NaN输入值将转换为正MAXNORM。

Returns

__nv_fp4x2_storage_t 值保存转换的结果。

__host__ __device__ __nv_fp4_storage_t __nv_cvt_halfraw_to_fp4(const __half_raw x, const __nv_fp4_interpretation_t fp4_interpretation, const enum cudaRoundMode rounding)

使用指定的舍入模式将输入的half精度x转换为请求类型的fp4，并对超出范围的值进行饱和处理。

将输入x转换为由fp4_interpretation参数指定类型的fp4格式，使用rounding参数指定的舍入模式。超出范围的大数值会饱和到相同符号的MAXNORM值。NaN输入值会转换为正的MAXNORM。

Returns

__nv_fp4_storage_t 值保存转换结果。

1.4.3. 类型定义

typedef __nv_fp8_storage_t __nv_fp4_storage_t: 用于存储fp4浮点数的8位unsignedinteger类型抽象。

typedef __nv_fp8_storage_t __nv_fp4x2_storage_t: 用于存储fp4浮点数对的8位unsignedinteger类型抽象。

typedef __nv_fp8x2_storage_t __nv_fp4x4_storage_t: 用于存储fp4浮点数四元组的16位unsignedinteger类型抽象。