3.9. FP8转换与数据移动

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

枚举

__nv_fp8_interpretation_t

列举将8位数值作为fp8类型时的可能解释方式。

__nv_saturation_t

列举了在执行向fp8目标类型进行窄化转换时适用的模式。

Functions

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_bfloat162raw_to_e8m0x2(const __nv_bfloat162_raw x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将一对bfloat16值转换为一对e8m0类型的缩放因子。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_bfloat16raw2_to_fp8x2(const __nv_bfloat162_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中两个打包在__nv_bfloat162_raw x内的nv_bfloat16精度数值，转换为两个指定类型的fp8值向量，采用最近偶数舍入法和指定的饱和模式。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_bfloat16raw_to_e8m0(const __nv_bfloat16_raw x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将输入的bfloat16转换为e8m0类型的缩放因子。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_bfloat16raw_to_fp8(const __nv_bfloat16_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入的nv_bfloat16精度x转换为指定类型的fp8格式，采用最近偶数舍入法和指定的饱和模式。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_double2_to_e8m0x2(const double2 x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将一对double值转换为e8m0类型的缩放因子对。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_double2_to_fp8x2(const double2 x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中两个打包在double2 x中的double精度数字，使用就近取偶舍入法和指定的饱和模式，转换为两个指定类型的fp8值向量。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_double_to_e8m0(const double x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将输入的double值转换为e8m0类型的缩放因子。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_double_to_fp8(const double x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

使用四舍五入到最近的偶数舍入方式和指定的饱和模式，将输入的double精度x转换为指定类型的fp8格式。

__host__ __device__ __nv_bfloat16_raw __nv_cvt_e8m0_to_bf16raw(const __nv_fp8_storage_t x)

将e8m0类型的输入缩放因子值转换为bfloat16。

__host__ __device__ __nv_bfloat162_raw __nv_cvt_e8m0x2_to_bf162raw(const __nv_fp8x2_storage_t x)

将输入的一对e8m0类型的缩放因子转换为一对bfloat16值。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_float2_to_e8m0x2(const float2 x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将一对float值转换为一对e8m0类型的缩放因子。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_float2_to_fp8x2(const float2 x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中的两个single精度数（打包在float2 x中）转换为两个fp8类型值的向量，使用最近偶数舍入方式和指定的饱和模式。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_float_to_e8m0(const float x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将输入的float值转换为e8m0类型的缩放因子。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_float_to_fp8(const float x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

使用就近偶数舍入和指定的饱和模式，将输入的single精度x转换为指定类型的fp8格式。

__host__ __device__ __half_raw __nv_cvt_fp8_to_halfraw(const __nv_fp8_storage_t x, const __nv_fp8_interpretation_t fp8_interpretation)

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

__host__ __device__ __half2_raw __nv_cvt_fp8x2_to_halfraw2(const __nv_fp8x2_storage_t x, const __nv_fp8_interpretation_t fp8_interpretation)

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

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_halfraw2_to_fp8x2(const __half2_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中两个打包在__half2_raw x内的半精度数，使用就近取偶舍入法和指定的饱和模式，转换为两个指定类型fp8值的向量。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_halfraw_to_fp8(const __half_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

使用四舍五入到最近的偶数舍入方式和指定的饱和模式，将输入的half精度x转换为请求类型的fp8格式。

__host__ __device__ __nv_fp8_e4m3::__nv_fp8_e4m3(const int val)

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

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

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

__host__ __device__ __nv_fp8_e4m3::__nv_fp8_e4m3(const __nv_bfloat16 f)

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

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

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

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

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

__nv_fp8_e4m3::__nv_fp8_e4m3()=default

默认构造函数。

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

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

__host__ __device__ __nv_fp8_e4m3::__nv_fp8_e4m3(const float f)

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

__host__ __device__ __nv_fp8_e4m3::__nv_fp8_e4m3(const __half f)

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

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

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

__host__ __device__ __nv_fp8_e4m3::__nv_fp8_e4m3(const double f)

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

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

从short int数据类型的构造函数，依赖__NV_SATFINITE处理超出范围值的行为。

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

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

__host__ __device__ __nv_fp8_e4m3::operator __half() const

转换为__half数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator __nv_bfloat16() const

转换为 __nv_bfloat16 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator bool() const

转换为bool数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator char() const

转换为实现定义的 char 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator double() const

转换为double数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator float() const

转换为float数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator int() const

转换为int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator long int() const

转换为 long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator long long int() const

转换为 long long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator short int() const

转换为short int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator signed char() const

转换为signed char数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator unsigned char() const

转换为 unsigned char 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator unsigned int() const

转换为unsigned int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator unsigned long int() const

转换为 unsigned long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator unsigned long long int() const

转换为 unsigned long long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e4m3::operator unsigned short int() const

转换为unsigned short int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::__nv_fp8_e5m2(const __half f)

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

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

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

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

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

__host__ __device__ __nv_fp8_e5m2::__nv_fp8_e5m2(const float f)

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

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

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

__nv_fp8_e5m2::__nv_fp8_e5m2()=default

默认构造函数。

__host__ __device__ __nv_fp8_e5m2::__nv_fp8_e5m2(const int val)

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

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

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

__host__ __device__ __nv_fp8_e5m2::__nv_fp8_e5m2(const double f)

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

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

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

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

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

__host__ __device__ __nv_fp8_e5m2::__nv_fp8_e5m2(const __nv_bfloat16 f)

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

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

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

__host__ __device__ __nv_fp8_e5m2::operator __half() const

转换为__half数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator __nv_bfloat16() const

转换为 __nv_bfloat16 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator bool() const

转换为bool数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator char() const

转换为实现定义的 char 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator double() const

转换为 double 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator float() const

转换为float数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator int() const

转换为int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator long int() const

转换为 long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator long long int() const

转换为 long long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator short int() const

转换为short int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator signed char() const

转换为signed char数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator unsigned char() const

转换为 unsigned char 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator unsigned int() const

转换为unsigned int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator unsigned long int() const

转换为 unsigned long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator unsigned long long int() const

转换为 unsigned long long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e5m2::operator unsigned short int() const

转换为unsigned short int数据类型的转换运算符。

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

从long int数据类型构造，依赖于cudaRoundZero舍入方式。

__host__ __device__ __nv_fp8_e8m0::__nv_fp8_e8m0(const int val)

从int数据类型构造，依赖于cudaRoundZero舍入方式。

__nv_fp8_e8m0::__nv_fp8_e8m0()=default

默认构造函数。

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

从unsigned int数据类型的构造函数，依赖于cudaRoundZero舍入方式。

__host__ __device__ __nv_fp8_e8m0::__nv_fp8_e8m0(const float f)

从float数据类型构造，对于大输入值依赖于__NV_SATFINITE的行为特性，并使用cudaRoundZero进行舍入。

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

从unsigned long long int数据类型构造，依赖于cudaRoundZero舍入方式。

__host__ __device__ __nv_fp8_e8m0::__nv_fp8_e8m0(const double f)

从double数据类型构造，对于大输入值依赖于__NV_SATFINITE行为，并使用cudaRoundZero进行舍入。

__host__ __device__ __nv_fp8_e8m0::__nv_fp8_e8m0(const __half f)

从__half数据类型构造，对于大输入值依赖于__NV_SATFINITE的行为，并使用cudaRoundZero进行舍入。

__host__ __device__ __nv_fp8_e8m0::__nv_fp8_e8m0(const __nv_bfloat16 f)

从__nv_bfloat16数据类型的构造函数，对于大输入值依赖__NV_SATFINITE行为，并使用cudaRoundZero进行舍入。

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

从unsigned long int数据类型构造，依赖于cudaRoundZero舍入方式。

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

从unsigned short int数据类型构造，依赖于cudaRoundZero舍入方式。

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

从long long int数据类型的构造函数，依赖于cudaRoundZero舍入方式。

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

从short int数据类型构造，依赖于cudaRoundZero舍入方式。

__host__ __device__ __nv_fp8_e8m0::operator __half() const

转换为__half数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator __nv_bfloat16() const

转换为 __nv_bfloat16 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator bool() const

转换为bool数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator char() const

转换为实现定义的 char 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator double() const

转换为double数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator float() const

转换为float数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator int() const

转换为int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator long int() const

转换为 long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator long long int() const

转换为 long long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator short int() const

转换为short int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator signed char() const

转换为signed char数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator unsigned char() const

转换为 unsigned char 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator unsigned int() const

转换为unsigned int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator unsigned long int() const

转换为 unsigned long int 数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator unsigned long long int() const

转换为unsigned long long int数据类型的转换运算符。

__host__ __device__ __nv_fp8_e8m0::operator unsigned short int() const

转换为unsigned short int数据类型的转换运算符。

__nv_fp8x2_e4m3::__nv_fp8x2_e4m3()=default

默认构造函数。

__host__ __device__ __nv_fp8x2_e4m3::__nv_fp8x2_e4m3(const __nv_bfloat162 f)

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

__host__ __device__ __nv_fp8x2_e4m3::__nv_fp8x2_e4m3(const double2 f)

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

__host__ __device__ __nv_fp8x2_e4m3::__nv_fp8x2_e4m3(const __half2 f)

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

__host__ __device__ __nv_fp8x2_e4m3::__nv_fp8x2_e4m3(const float2 f)

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

__host__ __device__ __nv_fp8x2_e4m3::operator __half2() const

转换为 __half2 数据类型的转换运算符。

__host__ __device__ __nv_fp8x2_e4m3::operator float2() const

转换为float2数据类型的转换运算符。

__host__ __device__ __nv_fp8x2_e5m2::__nv_fp8x2_e5m2(const __nv_bfloat162 f)

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

__host__ __device__ __nv_fp8x2_e5m2::__nv_fp8x2_e5m2(const double2 f)

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

__host__ __device__ __nv_fp8x2_e5m2::__nv_fp8x2_e5m2(const __half2 f)

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

__nv_fp8x2_e5m2::__nv_fp8x2_e5m2()=default

默认构造函数。

__host__ __device__ __nv_fp8x2_e5m2::__nv_fp8x2_e5m2(const float2 f)

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

__host__ __device__ __nv_fp8x2_e5m2::operator __half2() const

转换为__half2数据类型的转换运算符。

__host__ __device__ __nv_fp8x2_e5m2::operator float2() const

转换为float2数据类型的转换运算符。

__host__ __device__ __nv_fp8x2_e8m0::__nv_fp8x2_e8m0(const __half2 f)

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

__host__ __device__ __nv_fp8x2_e8m0::__nv_fp8x2_e8m0(const float2 f)

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

__host__ __device__ __nv_fp8x2_e8m0::__nv_fp8x2_e8m0(const __nv_bfloat162 f)

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

__host__ __device__ __nv_fp8x2_e8m0::__nv_fp8x2_e8m0(const double2 f)

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

__nv_fp8x2_e8m0::__nv_fp8x2_e8m0()=default

默认构造函数。

__host__ __device__ __nv_fp8x2_e8m0::operator __half2() const

转换为 __half2 数据类型的转换运算符。

__host__ __device__ __nv_fp8x2_e8m0::operator __nv_bfloat162() const

转换为 __nv_bfloat162 数据类型的转换运算符。

__host__ __device__ __nv_fp8x2_e8m0::operator float2() const

转换为float2数据类型的转换运算符。

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

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

__host__ __device__ __nv_fp8x4_e4m3::__nv_fp8x4_e4m3(const double4 f)

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

__nv_fp8x4_e4m3::__nv_fp8x4_e4m3()=default

默认构造函数。

__host__ __device__ __nv_fp8x4_e4m3::__nv_fp8x4_e4m3(const float4 f)

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

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

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

__host__ __device__ __nv_fp8x4_e4m3::operator float4() const

转换为float4向量数据类型的转换运算符。

__nv_fp8x4_e5m2::__nv_fp8x4_e5m2()=default

默认构造函数。

__host__ __device__ __nv_fp8x4_e5m2::__nv_fp8x4_e5m2(const double4 f)

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

__host__ __device__ __nv_fp8x4_e5m2::__nv_fp8x4_e5m2(const float4 f)

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

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

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

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

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

__host__ __device__ __nv_fp8x4_e5m2::operator float4() const

转换为float4向量数据类型的转换运算符。

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

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

__host__ __device__ __nv_fp8x4_e8m0::__nv_fp8x4_e8m0(const float4 f)

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

__host__ __device__ __nv_fp8x4_e8m0::__nv_fp8x4_e8m0(const double4 f)

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

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

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

__nv_fp8x4_e8m0::__nv_fp8x4_e8m0()=default

默认构造函数。

__host__ __device__ __nv_fp8x4_e8m0::operator float4() const

转换为float4向量数据类型的转换运算符。

类型定义

__nv_fp8_storage_t

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

__nv_fp8x2_storage_t

用于存储fp8浮点数对的16位unsigned integer类型抽象。

__nv_fp8x4_storage_t

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

3.9.1. 枚举

enum __nv_fp8_interpretation_t

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

取值：

enumerator __NV_E4M3

代表fp8类型的e4m3数值。

enumerator __NV_E5M2

代表fp8格式的e5m2类型数值。

enum __nv_saturation_t

列举了在执行向fp8目标类型进行窄化转换时适用的模式。

取值：

enumerator __NV_NOSAT

表示当转换结果导致舍入值超出目标类型范围时，不执行对有限值的饱和处理。

注意：对于e4m3类型的fp8格式，大于目标格式可表示有限数最大值的结果将变为NaN。

enumerator __NV_SATFINITE

表示输入值大于目标格式可表示的最大有限数MAXNORM时，将舍入为与输入值同符号的MAXNORM值。

3.9.2. 函数

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_bfloat162raw_to_e8m0x2(const __nv_bfloat162_raw x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将一对bfloat16值转换为一对e8m0类型的缩放因子。

另请参阅

__nv_cvt_bfloat16raw_to_e8m0() 转换的详细信息。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_bfloat16raw2_to_fp8x2(const __nv_bfloat162_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中两个nv_bfloat16精度的数字（打包在__nv_bfloat162_raw x中）转换为两个fp8类型值的向量，转换过程采用就近偶数舍入法和指定的饱和模式。

将输入向量 x 转换为由 fp8_interpretation 参数指定类型的两个 fp8 值组成的向量，使用由 saturate 参数指定的就近偶数舍入和饱和模式。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_bfloat16raw_to_e8m0(const __nv_bfloat16_raw x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将输入的bfloat16转换为e8m0类型的缩放因子。

输入数字的绝对值会根据rounding参数指定的方向四舍五入到最接近的2的幂次方。若舍入结果小于可表示的最小目标格式数值2^-127，则会被截断为2^-127。若结果大于可表示的最大目标格式数值2^127，当saturate等于__NV_SATFINITE时会被截断为2^127，否则将转换为NaN。输入为NaN时将输出NaN，在目标格式中编码为0xFF。

Returns

• __nv_fp8_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_bfloat16raw_to_fp8(const __nv_bfloat16_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入的nv_bfloat16精度x转换为请求类型的fp8格式，采用最近偶数舍入法和请求的饱和模式。

将输入x转换为由fp8_interpretation参数指定类型的fp8格式，使用就近取偶舍入方式，并根据saturate参数指定的饱和模式进行处理。

Returns

• __nv_fp8_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_double2_to_e8m0x2(const double2 x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将一对double值转换为e8m0类型的缩放因子对。

另请参阅

__nv_cvt_bfloat16raw_to_e8m0() 转换的详细信息。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_double2_to_fp8x2(const double2 x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中两个打包在double2 x的双精度数字转换为两个fp8类型值的向量，使用就近取偶舍入和指定的饱和模式。

将输入向量 x 转换为由 fp8_interpretation 参数指定类型的两个 fp8 值组成的向量，使用由 saturate 参数指定的就近偶数舍入和饱和模式。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_double_to_e8m0(const double x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将输入的double值转换为e8m0类型的缩放因子。

另请参阅

__nv_cvt_bfloat16raw_to_e8m0() 转换的详细信息。

Returns

• __nv_fp8_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_double_to_fp8(const double x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

使用四舍五入到最近的偶数舍入方式和指定的饱和模式，将输入的double精度x转换为指定类型的fp8格式。

将输入x转换为由fp8_interpretation参数指定类型的fp8格式，使用就近取偶舍入方式，并根据saturate参数指定的饱和模式进行处理。

Returns

• __nv_fp8_storage_t 值保存转换结果。

__host__ __device__ __nv_bfloat16_raw __nv_cvt_e8m0_to_bf16raw(const __nv_fp8_storage_t x)

将e8m0类型的输入缩放因子值转换为bfloat16。

输入比例是精确的2的幂次方或一个NaN值，同样可以在目标格式中表示。

Returns

• __nv_bfloat16_raw 值保存转换结果。

__host__ __device__ __nv_bfloat162_raw __nv_cvt_e8m0x2_to_bf162raw(const __nv_fp8x2_storage_t x)

将输入的e8m0类型缩放因子对转换为bfloat16值对。

Returns

• __nv_bfloat162_raw 值保存转换结果。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_float2_to_e8m0x2(const float2 x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将一对float值转换为一对e8m0类型的缩放因子。

另请参阅

__nv_cvt_bfloat16raw_to_e8m0() 转换的详细信息。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_float2_to_fp8x2(const float2 x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中的两个single精度数（打包在float2 x中）转换为两个fp8类型值的向量，采用就近偶数舍入方式及指定的饱和模式。

将输入向量 x 转换为由 fp8_interpretation 参数指定类型的两个 fp8 值组成的向量，使用由 saturate 参数指定的就近偶数舍入和饱和模式。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_float_to_e8m0(const float x, const __nv_saturation_t saturate, const enum cudaRoundMode rounding)

将输入的float值转换为e8m0类型的缩放因子。

另请参阅

__nv_cvt_bfloat16raw_to_e8m0() 转换的详细信息。

Returns

• __nv_fp8_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_float_to_fp8(const float x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

使用就近偶数舍入和指定的饱和模式，将输入的single精度x转换为指定类型的fp8格式。

将输入x转换为由fp8_interpretation参数指定类型的fp8格式，使用就近取偶舍入方式，并根据saturate参数指定的饱和模式进行处理。

Returns

• __nv_fp8_storage_t 值保存转换结果。

__host__ __device__ __half_raw __nv_cvt_fp8_to_halfraw(const __nv_fp8_storage_t x, const __nv_fp8_interpretation_t fp8_interpretation)

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

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

Returns

• __half_raw 值保存转换结果。

__host__ __device__ __half2_raw __nv_cvt_fp8x2_to_halfraw2(const __nv_fp8x2_storage_t x, const __nv_fp8_interpretation_t fp8_interpretation)

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

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

Returns

• __half2_raw 值保存转换结果。

__host__ __device__ __nv_fp8x2_storage_t __nv_cvt_halfraw2_to_fp8x2(const __half2_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

将输入向量中的两个half精度数（打包在__half2_raw x中）转换为两个fp8类型值的向量，使用就近取偶舍入方式和指定的饱和模式。

将输入向量 x 转换为由 fp8_interpretation 参数指定类型的两个 fp8 值组成的向量，使用由 saturate 参数指定的就近偶数舍入和饱和模式。

Returns

• __nv_fp8x2_storage_t 值保存转换结果。

__host__ __device__ __nv_fp8_storage_t __nv_cvt_halfraw_to_fp8(const __half_raw x, const __nv_saturation_t saturate, const __nv_fp8_interpretation_t fp8_interpretation)

使用最近偶数舍入法和指定的饱和模式，将输入的half精度x转换为指定类型的fp8格式。

将输入x转换为由fp8_interpretation参数指定类型的fp8格式，使用就近取偶舍入方式，并根据saturate参数指定的饱和模式进行处理。

Returns

• __nv_fp8_storage_t 值保存转换结果。

3.9.3. 类型定义

typedef unsigned char __nv_fp8_storage_t

用于存储fp8浮点数的8位unsigned integer类型抽象。

typedef unsigned short int __nv_fp8x2_storage_t

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

typedef unsigned int __nv_fp8x4_storage_t

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