__bfloat162half_ru_sat

Applicability

Product

Supported

Atlas 350 Accelerator Card

Atlas A3 training product/Atlas A3 inference product

x

Atlas A2 training product/Atlas A2 inference product

x

Atlas 200I/500 A2 inference product

x

Atlas inference product AI Core

x

Atlas inference product Vector Core

x

Atlas training product

x

Function Usage

In saturation mode, converts bfloat16 data to the half type in line with the CAST_CEIL mode and returns the converted value.

Prototype

1
inline half __bfloat162half_ru_sat(const bfloat16_t x)

Parameters

Table 1 Parameters

Parameter

Input/Output

Description

x

Input

Source operand.

Returns

The half data converted from the input in saturation mode in line with the CAST_CEIL mode.

Restrictions

Before using this API, set bit[60] of the CTRL register to 0. Otherwise, the saturation mode does not take effect. For details, see Methods of Controlling Saturation Behaviors.

For SIMT programming, this API is not supported.

Header File to Be Included

To use this API, the simt_api/asc_bf16.h header file must be included.

1
#include "simt_api/asc_bf16.h"

Examples

For SIMD and SIMT programming:
1
2
3
4
5
__simt_vf__ __launch_bounds__(1024) inline void kernel__bfloat162half_ru_sat(__gm__ half* dst, __gm__ bfloat16_t* x)
{
    int idx = threadIdx.x + blockIdx.x * blockDim.x;
    dst[idx] = __bfloat162half_ru_sat(x[idx]);
}