2.7. FP6转换与数据移动

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

枚举

__nv_fp6_interpretation_t

列举将8位值作为fp6类型时的可能解释。

Functions

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_bfloat16raw2_to_fp6x2(const __nv_bfloat162_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6_storage_t __nv_cvt_bfloat16raw_to_fp6(const __nv_bfloat16_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_double2_to_fp6x2(const double2 x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6_storage_t __nv_cvt_double_to_fp6(const double x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_float2_to_fp6x2(const float2 x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6_storage_t __nv_cvt_float_to_fp6(const float x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __half_raw __nv_cvt_fp6_to_halfraw(const __nv_fp6_storage_t x, const __nv_fp6_interpretation_t fp6_interpretation)

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

__host__ __device__ __half2_raw __nv_cvt_fp6x2_to_halfraw2(const __nv_fp6x2_storage_t x, const __nv_fp6_interpretation_t fp6_interpretation)

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

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_halfraw2_to_fp6x2(const __half2_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6_storage_t __nv_cvt_halfraw_to_fp6(const __half_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3()

默认构造函数。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const float f)

从float数据类型构造，对于超出范围的值依赖__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const long long int val)

从long long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const long int val)

从long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const unsigned short int val)

从unsigned short int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const double f)

从double数据类型构造，对于超出范围的值依赖于__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const unsigned long long int val)

从unsigned long long int数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const short int val)

从short int数据类型的构造函数。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const __nv_bfloat16 f)

从__nv_bfloat16数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const int val)

从int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const __half f)

从__half数据类型构造，依赖__NV_SATFINITE处理超出范围值的行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const unsigned int val)

从unsigned int数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e2m3::__nv_fp6_e2m3(const unsigned long int val)

从unsigned long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const long long int val)

从long long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const __half f)

从__half数据类型构造，依赖__NV_SATFINITE处理超出范围值的行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const int val)

从int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const float f)

从float数据类型构造，对于超出范围的值依赖__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const short int val)

从short int数据类型的构造函数。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const long int val)

从long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const unsigned long long int val)

从unsigned long long int数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const unsigned short int val)

从unsigned short int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const __nv_bfloat16 f)

从__nv_bfloat16数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const double f)

从double数据类型构造，对于超出范围的值依赖于__NV_SATFINITE行为，并使用cudaRoundNearest舍入模式。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const unsigned int val)

从unsigned int数据类型的构造函数，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2()

默认构造函数。

__host__ __device__ __nv_fp6_e3m2::__nv_fp6_e3m2(const unsigned long int val)

从unsigned long int数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e2m3::__nv_fp6x2_e2m3(const __nv_bfloat162 f)

从__nv_bfloat162数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e2m3::__nv_fp6x2_e2m3(const float2 f)

从float2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e2m3::__nv_fp6x2_e2m3(const double2 f)

从double2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e2m3::__nv_fp6x2_e2m3(const __half2 f)

从__half2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e2m3::__nv_fp6x2_e2m3()

默认构造函数。

__host__ __device__ __nv_fp6x2_e3m2::__nv_fp6x2_e3m2(const __half2 f)

从__half2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e3m2::__nv_fp6x2_e3m2(const float2 f)

从float2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e3m2::__nv_fp6x2_e3m2(const __nv_bfloat162 f)

从__nv_bfloat162数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e3m2::__nv_fp6x2_e3m2(const double2 f)

从double2数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x2_e3m2::__nv_fp6x2_e3m2()

默认构造函数。

__host__ __device__ __nv_fp6x4_e2m3::__nv_fp6x4_e2m3(const __half2 flo, const __half2 fhi)

从一对__half2数据类型值构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e2m3::__nv_fp6x4_e2m3(const float4 f)

从float4向量数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e2m3::__nv_fp6x4_e2m3(const double4 f)

从double4向量数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e2m3::__nv_fp6x4_e2m3(const __nv_bfloat162 flo, const __nv_bfloat162 fhi)

从一对__nv_bfloat162数据类型值构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e2m3::__nv_fp6x4_e2m3()

默认构造函数。

__host__ __device__ __nv_fp6x4_e3m2::__nv_fp6x4_e3m2()

默认构造函数。

__host__ __device__ __nv_fp6x4_e3m2::__nv_fp6x4_e3m2(const __half2 flo, const __half2 fhi)

从一对__half2数据类型值构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e3m2::__nv_fp6x4_e3m2(const float4 f)

从float4向量数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e3m2::__nv_fp6x4_e3m2(const __nv_bfloat162 flo, const __nv_bfloat162 fhi)

从一对__nv_bfloat162数据类型值构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

__host__ __device__ __nv_fp6x4_e3m2::__nv_fp6x4_e3m2(const double4 f)

从double4向量数据类型构造，对于超出范围的值依赖于__NV_SATFINITE的行为。

类型定义

__nv_fp6_storage_t

用于fp6浮点数存储的8位unsignedinteger类型抽象。

__nv_fp6x2_storage_t

16位unsignedinteger类型抽象，用于存储fp6浮点数对。

__nv_fp6x4_storage_t

用于存储fp6浮点数四元组的32位unsigned integer类型抽象。

2.7.1. 枚举

enum __nv_fp6_interpretation_t

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

取值：

enumerator __NV_E2M3

表示fp6类型的e2m3数值。

enumerator __NV_E3M2

代表fp6类型的e3m2数值。

2.7.2. 函数

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_bfloat16raw2_to_fp6x2(const __nv_bfloat162_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp6_storage_t __nv_cvt_bfloat16raw_to_fp6(const __nv_bfloat16_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6_storage_t 值保存转换结果。

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_double2_to_fp6x2(const double2 x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp6_storage_t __nv_cvt_double_to_fp6(const double x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6_storage_t 值保存转换结果。

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_float2_to_fp6x2(const float2 x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp6_storage_t __nv_cvt_float_to_fp6(const float x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6_storage_t 值保存转换结果。

__host__ __device__ __half_raw __nv_cvt_fp6_to_halfraw(const __nv_fp6_storage_t x, const __nv_fp6_interpretation_t fp6_interpretation)

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

将输入参数x从fp6类型（由fp6_interpretation参数指定具体格式）转换为half精度。

Returns

• __half_raw 值保存转换结果。

__host__ __device__ __half2_raw __nv_cvt_fp6x2_to_halfraw2(const __nv_fp6x2_storage_t x, const __nv_fp6_interpretation_t fp6_interpretation)

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

将输入向量 x（类型为 fp6，具体格式由参数 fp6_interpretation 指定）转换为两个 half 精度的值向量，并以 __half2_raw 结构体形式返回。

Returns

• __half2_raw 值保存转换结果。

__host__ __device__ __nv_fp6x2_storage_t __nv_cvt_halfraw2_to_fp6x2(const __half2_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp6_storage_t __nv_cvt_halfraw_to_fp6(const __half_raw x, const __nv_fp6_interpretation_t fp6_interpretation, const enum cudaRoundMode rounding)

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

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

Returns

• __nv_fp6_storage_t 值保存转换结果。

2.7.3. 类型定义

typedef __nv_fp8_storage_t __nv_fp6_storage_t

用于fp6浮点数存储的8位unsignedinteger类型抽象。

typedef __nv_fp8x2_storage_t __nv_fp6x2_storage_t

用于存储fp6浮点数对的16位无符号整型抽象类型。

typedef __nv_fp8x4_storage_t __nv_fp6x4_storage_t

用于存储fp6浮点数四元组的32位unsigned integer类型抽象。