__float2half_ro_sat

【免费下载链接】asc-devkit 本项目是CANN 推出的昇腾AI处理器专用的算子程序开发语言,原生支持C和C++标准规范,主要由类库和语言扩展层构成,提供多层级API,满足多维场景算子开发诉求。 【免费下载链接】asc-devkit 项目地址: https://gitcode.com/cann/asc-devkit

产品支持情况

  • Ascend 950PR/Ascend 950DT:支持
  • Atlas A3 训练系列产品/Atlas A3 推理系列产品:不支持
  • Atlas A2 训练系列产品/Atlas A2 推理系列产品:不支持
  • Atlas 200I/500 A2 推理产品:不支持
  • Atlas 推理系列产品AI Core:不支持
  • Atlas 推理系列产品Vector Core:不支持
  • Atlas 训练系列产品:不支持

功能说明

将浮点数转换为半精度浮点数,并遵循CAST_ODD模式,返回转换后的值。

函数原型

inline half __float2half_ro_sat(const float x)

参数说明

表1 参数说明

参数名 输入/输出 描述
x 输入 源操作数。

返回值说明

将输入遵循CAST_ODD模式转换成的半精度浮点数。本接口受全局饱和寄存器的影响,特殊值如下:

x值 非饱和模式返回值 饱和模式返回值
±0 0 0
nan nan 0
inf inf ASCRT_MAX_NORMAL_FP16
-inf -inf -ASCRT_MAX_NORMAL_FP16
x大于且接近ASCRT_MAX_NORMAL_FP16 ASCRT_MAX_NORMAL_FP16 ASCRT_MAX_NORMAL_FP16
x小于且接近-ASCRT_MAX_NORMAL_FP16 -ASCRT_MAX_NORMAL_FP16 -ASCRT_MAX_NORMAL_FP16
x远大于ASCRT_MAX_NORMAL_FP16 inf ASCRT_MAX_NORMAL_FP16
x远小于-ASCRT_MAX_NORMAL_FP16 -inf -ASCRT_MAX_NORMAL_FP16

约束说明

使用此接口前需将CTRL[60]寄存器设置为0,否则饱和模式不生效。设置方式请参见控制饱和行为的方式

SIMT编程场景由于无法设置CTRL寄存器,本接口的饱和模式不生效。

需要包含的头文件

使用该接口需要包含"simt_api/asc_fp16.h"头文件。

#include "simt_api/asc_fp16.h"

调用示例

  • SIMD与SIMT混合编程场景:

    __simt_vf__ __launch_bounds__(1024) inline void kernel__float2half_ro_sat(__gm__ half* dst, __gm__ float* x)
    {
        int idx = threadIdx.x + blockIdx.x * blockDim.x;
        dst[idx] = __float2half_ro_sat(x[idx]);
    }
    

【免费下载链接】asc-devkit 本项目是CANN 推出的昇腾AI处理器专用的算子程序开发语言,原生支持C和C++标准规范,主要由类库和语言扩展层构成,提供多层级API,满足多维场景算子开发诉求。 【免费下载链接】asc-devkit 项目地址: https://gitcode.com/cann/asc-devkit

Logo

Agent 垂直技术社区,欢迎活跃、内容共建。

更多推荐