14. SIMD Intrinsics

本节介绍仅在设备代码中支持的 SIMD 内在函数。

要使用这些函数，您无需在程序中包含任何额外的头文件。

函数

__device__ unsigned int __vabs2(unsigned int a)

计算每半字绝对值：|a|。

__device__ unsigned int __vabs4(unsigned int a)

计算每字节绝对值：|a|。

__device__ unsigned int __vabsdiffs2(unsigned int a, unsigned int b)

计算每半字有符号整数的绝对差值：|a - b|。

__device__ unsigned int __vabsdiffs4(unsigned int a, unsigned int b)

计算每字节有符号整数的绝对差值：|a - b|。

__device__ unsigned int __vabsdiffu2(unsigned int a, unsigned int b)

计算每半字无符号整数的绝对差值：|a - b|。

__device__ unsigned int __vabsdiffu4(unsigned int a, unsigned int b)

计算每字节无符号整数的绝对差值：|a - b|。

__device__ unsigned int __vabsss2(unsigned int a)

计算每半字带符号饱和的绝对值：|a|。

__device__ unsigned int __vabsss4(unsigned int a)

计算每字节带符号饱和的绝对值：|a|。

__device__ unsigned int __vadd2(unsigned int a, unsigned int b)

执行每半字（无）符号加法，带环绕：a + b。

__device__ unsigned int __vadd4(unsigned int a, unsigned int b)

执行每字节（无）符号加法：a + b。

__device__ unsigned int __vaddss2(unsigned int a, unsigned int b)

执行每半字带符号饱和的加法：a + b。

__device__ unsigned int __vaddss4(unsigned int a, unsigned int b)

执行每字节带符号饱和的加法：a + b。

__device__ unsigned int __vaddus2(unsigned int a, unsigned int b)

执行每半字带无符号饱和的加法：a + b。

__device__ unsigned int __vaddus4(unsigned int a, unsigned int b)

执行每字节带无符号饱和的加法：a + b。

__device__ unsigned int __vavgs2(unsigned int a, unsigned int b)

执行每半字有符号舍入平均值计算。

__device__ unsigned int __vavgs4(unsigned int a, unsigned int b)

计算每字节有符号舍入平均值。

__device__ unsigned int __vavgu2(unsigned int a, unsigned int b)

执行每半字无符号舍入平均值计算。

__device__ unsigned int __vavgu4(unsigned int a, unsigned int b)

执行每字节无符号舍入平均值。

__device__ unsigned int __vcmpeq2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：a == b ? 0xffff : 0。

__device__ unsigned int __vcmpeq4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：a == b ? 0xff : 0。

__device__ unsigned int __vcmpges2(unsigned int a, unsigned int b)

执行每半字有符号比较：a >= b ? 0xffff : 0。

__device__ unsigned int __vcmpges4(unsigned int a, unsigned int b)

执行每字节有符号比较：a >= b ? 0xff : 0。

__device__ unsigned int __vcmpgeu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a >= b ? 0xffff : 0。

__device__ unsigned int __vcmpgeu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a >= b ? 0xff : 0。

__device__ unsigned int __vcmpgts2(unsigned int a, unsigned int b)

执行每半字有符号比较：a > b ? 0xffff : 0。

__device__ unsigned int __vcmpgts4(unsigned int a, unsigned int b)

执行每字节有符号比较：a > b ? 0xff : 0。

__device__ unsigned int __vcmpgtu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a > b ? 0xffff : 0。

__device__ unsigned int __vcmpgtu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a > b ? 0xff : 0。

__device__ unsigned int __vcmples2(unsigned int a, unsigned int b)

执行每半字有符号比较：a <= b ? 0xffff : 0。

__device__ unsigned int __vcmples4(unsigned int a, unsigned int b)

执行每字节有符号比较：a <= b ? 0xff : 0。

__device__ unsigned int __vcmpleu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a <= b ? 0xffff : 0。

__device__ unsigned int __vcmpleu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a <= b ? 0xff : 0。

__device__ unsigned int __vcmplts2(unsigned int a, unsigned int b)

执行每半字有符号比较：a < b ? 0xffff : 0。

__device__ unsigned int __vcmplts4(unsigned int a, unsigned int b)

执行每字节有符号比较：a < b ? 0xff : 0。

__device__ unsigned int __vcmpltu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a < b ? 0xffff : 0。

__device__ unsigned int __vcmpltu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a < b ? 0xff : 0。

__device__ unsigned int __vcmpne2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：a != b ? 0xffff : 0。

__device__ unsigned int __vcmpne4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：a != b ? 0xff : 0。

__device__ unsigned int __vhaddu2(unsigned int a, unsigned int b)

执行每半字无符号平均值计算。

__device__ unsigned int __vhaddu4(unsigned int a, unsigned int b)

计算每字节无符号平均值。

__host__ __device__ unsigned int __viaddmax_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(a + b, c)

__host__ __device__ unsigned int __viaddmax_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(a + b, c), 0)

__host__ __device__ int __viaddmax_s32(const int a, const int b, const int c)

计算 max(a + b, c)

__host__ __device__ int __viaddmax_s32_relu(const int a, const int b, const int c)

计算 max(max(a + b, c), 0)

__host__ __device__ unsigned int __viaddmax_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(a + b, c)

__host__ __device__ unsigned int __viaddmax_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 max(a + b, c)

__host__ __device__ unsigned int __viaddmin_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(a + b, c)

__host__ __device__ unsigned int __viaddmin_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(min(a + b, c), 0)

__host__ __device__ int __viaddmin_s32(const int a, const int b, const int c)

计算 min(a + b, c)

__host__ __device__ int __viaddmin_s32_relu(const int a, const int b, const int c)

计算 max(min(a + b, c), 0)

__host__ __device__ unsigned int __viaddmin_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(a + b, c)

__host__ __device__ unsigned int __viaddmin_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 min(a + b, c)

__host__ __device__ unsigned int __vibmax_s16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 max(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a >= b) 的结果。

__host__ __device__ int __vibmax_s32(const int a, const int b, bool *const pred)

计算 max(a, b)，同时将 pred 指向的值设置为 (a >= b)。

__host__ __device__ unsigned int __vibmax_u16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 max(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a >= b) 的结果。

__host__ __device__ unsigned int __vibmax_u32(const unsigned int a, const unsigned int b, bool *const pred)

计算 max(a, b)，同时将 pred 指向的值设置为 (a >= b)。

__host__ __device__ unsigned int __vibmin_s16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 min(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a <= b) 的结果。

__host__ __device__ int __vibmin_s32(const int a, const int b, bool *const pred)

计算 min(a, b)，同时将 pred 指向的值设置为 (a <= b)。

__host__ __device__ unsigned int __vibmin_u16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 min(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a <= b) 的结果。

__host__ __device__ unsigned int __vibmin_u32(const unsigned int a, const unsigned int b, bool *const pred)

计算 min(a, b)，同时将 pred 指向的值设置为 (a <= b)。

__host__ __device__ unsigned int __vimax3_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(a, b), c)

__host__ __device__ unsigned int __vimax3_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(max(a, b), c), 0)

__host__ __device__ int __vimax3_s32(const int a, const int b, const int c)

计算 max(max(a, b), c)

__host__ __device__ int __vimax3_s32_relu(const int a, const int b, const int c)

计算 max(max(max(a, b), c), 0)

__host__ __device__ unsigned int __vimax3_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(a, b), c)

__host__ __device__ unsigned int __vimax3_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 max(max(a, b), c)

__host__ __device__ unsigned int __vimax_s16x2_relu(const unsigned int a, const unsigned int b)

执行每半字 max(max(a, b), 0)

__host__ __device__ int __vimax_s32_relu(const int a, const int b)

计算 max(max(a, b), 0)

__host__ __device__ unsigned int __vimin3_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(min(a, b), c)

__host__ __device__ unsigned int __vimin3_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(min(min(a, b), c), 0)

__host__ __device__ int __vimin3_s32(const int a, const int b, const int c)

计算 min(min(a, b), c)

__host__ __device__ int __vimin3_s32_relu(const int a, const int b, const int c)

计算 max(min(min(a, b), c), 0)

__host__ __device__ unsigned int __vimin3_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(min(a, b), c)

__host__ __device__ unsigned int __vimin3_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 min(min(a, b), c)

__host__ __device__ unsigned int __vimin_s16x2_relu(const unsigned int a, const unsigned int b)

执行每半字 max(min(a, b), 0)

__host__ __device__ int __vimin_s32_relu(const int a, const int b)

计算 max(min(a, b), 0)

__device__ unsigned int __vmaxs2(unsigned int a, unsigned int b)

执行每半字有符号最大值计算。

__device__ unsigned int __vmaxs4(unsigned int a, unsigned int b)

计算每字节有符号最大值。

__device__ unsigned int __vmaxu2(unsigned int a, unsigned int b)

执行每半字无符号最大值计算。

__device__ unsigned int __vmaxu4(unsigned int a, unsigned int b)

计算每字节无符号最大值。

__device__ unsigned int __vmins2(unsigned int a, unsigned int b)

执行每半字有符号最小值计算。

__device__ unsigned int __vmins4(unsigned int a, unsigned int b)

计算每字节有符号最小值。

__device__ unsigned int __vminu2(unsigned int a, unsigned int b)

执行每半字无符号最小值计算。

__device__ unsigned int __vminu4(unsigned int a, unsigned int b)

计算每字节无符号最小值。

__device__ unsigned int __vneg2(unsigned int a)

计算每半字取反。

__device__ unsigned int __vneg4(unsigned int a)

执行每字节取反。

__device__ unsigned int __vnegss2(unsigned int a)

计算每半字带符号饱和的取反。

__device__ unsigned int __vnegss4(unsigned int a)

执行每字节带符号饱和的取反。

__device__ unsigned int __vsads2(unsigned int a, unsigned int b)

执行每半字有符号绝对差值之和。

__device__ unsigned int __vsads4(unsigned int a, unsigned int b)

计算每字节有符号绝对差值之和。

__device__ unsigned int __vsadu2(unsigned int a, unsigned int b)

计算每半字无符号绝对差值之和。

__device__ unsigned int __vsadu4(unsigned int a, unsigned int b)

计算每字节无符号绝对差值之和。

__device__ unsigned int __vseteq2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：如果两个部分都相等，则返回 1。

__device__ unsigned int __vseteq4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：如果所有 4 对都相等，则返回 1。

__device__ unsigned int __vsetges2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都大于或等于，则返回 1。

__device__ unsigned int __vsetges4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都大于或等于，则返回 1。

__device__ unsigned int __vsetgeu2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都大于或等于，则返回 1。

__device__ unsigned int __vsetgeu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都大于或等于，则返回 1。

__device__ unsigned int __vsetgts2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都大于，则返回 1。

__device__ unsigned int __vsetgts4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都大于，则返回 1。

__device__ unsigned int __vsetgtu2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都大于，则返回 1。

__device__ unsigned int __vsetgtu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都大于，则返回 1。

__device__ unsigned int __vsetles2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都小于或等于，则返回 1。

__device__ unsigned int __vsetles4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都小于或等于，则返回 1。

__device__ unsigned int __vsetleu2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都小于或等于，则返回 1。

__device__ unsigned int __vsetleu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都小于或等于，则返回 1。

__device__ unsigned int __vsetlts2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都小于，则返回 1。

__device__ unsigned int __vsetlts4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都小于，则返回 1。

__device__ unsigned int __vsetltu2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都小于，则返回 1。

__device__ unsigned int __vsetltu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都小于，则返回 1。

__device__ unsigned int __vsetne2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：如果两个部分都不相等，则返回 1。

__device__ unsigned int __vsetne4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：如果所有 4 对都不相等，则返回 1。

__device__ unsigned int __vsub2(unsigned int a, unsigned int b)

执行每半字（无）符号减法，带环绕：a - b。

__device__ unsigned int __vsub4(unsigned int a, unsigned int b)

执行每字节减法：a - b。

__device__ unsigned int __vsubss2(unsigned int a, unsigned int b)

执行每半字（无）符号减法，带符号饱和：a - b。

__device__ unsigned int __vsubss4(unsigned int a, unsigned int b)

执行每字节带符号饱和的减法：a - b。

__device__ unsigned int __vsubus2(unsigned int a, unsigned int b)

执行每半字带无符号饱和的减法：a - b。

__device__ unsigned int __vsubus4(unsigned int a, unsigned int b)

执行每字节带无符号饱和的减法：a - b。

14.1. 函数

__device__ unsigned int __vabs2(unsigned int a)

计算每半字绝对值：|a|。

将参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后计算每个部分的绝对值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabs4(unsigned int a)

计算每字节绝对值：|a|。

按字节拆分参数。计算每个字节的绝对值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabsdiffs2(unsigned int a, unsigned int b)

计算每半字有符号整数的绝对差值：|a - b|。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，函数计算绝对差值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabsdiffs4(unsigned int a, unsigned int b)

计算每字节有符号整数的绝对差值：|a - b|。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，函数计算绝对差值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabsdiffu2(unsigned int a, unsigned int b)

计算每半字无符号整数的绝对差值：|a - b|。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，函数计算绝对差值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabsdiffu4(unsigned int a, unsigned int b)

计算每字节无符号整数的绝对差值：|a - b|。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，函数计算绝对差值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabsss2(unsigned int a)

计算每半字带符号饱和的绝对值：|a|。

将参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后计算每个部分的带符号饱和绝对值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vabsss4(unsigned int a)

计算每字节带符号饱和的绝对值：|a|。

将参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节，然后计算每个部分的带符号饱和绝对值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vadd2(unsigned int a, unsigned int b)

执行每半字（无）符号加法，带环绕：a + b。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后对对应的部分执行无符号加法。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vadd4(unsigned int a, unsigned int b)

执行每字节（无）符号加法：a + b。

将 'a' 拆分为 4 个字节，然后将这些字节中的每一个与 'b' 中对应的字节执行无符号加法，忽略溢出。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vaddss2(unsigned int a, unsigned int b)

执行每半字带符号饱和的加法：a + b。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后对对应的部分执行带符号饱和加法。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vaddss4(unsigned int a, unsigned int b)

执行每字节带符号饱和的加法：a + b。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节，然后对对应的部分执行带符号饱和加法。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vaddus2(unsigned int a, unsigned int b)

执行每半字带无符号饱和的加法：a + b。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后对对应的部分执行无符号饱和加法。

返回值

返回计算后的值。

__device__ unsigned int __vaddus4(unsigned int a, unsigned int b)

执行每字节带无符号饱和的加法：a + b。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节，然后对对应的部分执行无符号饱和加法。

返回值

返回计算后的值。

__device__ unsigned int __vavgs2(unsigned int a, unsigned int b)

执行每半字有符号舍入平均值计算。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后计算对应部分的带符号四舍五入平均值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vavgs4(unsigned int a, unsigned int b)

计算每字节有符号舍入平均值。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。然后计算对应部分的带符号四舍五入平均值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vavgu2(unsigned int a, unsigned int b)

执行每半字无符号舍入平均值计算。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节，然后计算对应部分的无符号四舍五入平均值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vavgu4(unsigned int a, unsigned int b)

执行每字节无符号舍入平均值。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。然后计算对应部分的无符号四舍五入平均值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vcmpeq2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：a == b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果它们相等，则结果为 ffff，否则为 0000。例如，__vcmpeq2(0x1234aba5, 0x1234aba6) 返回 0xffff0000。

返回值

返回 0xffff 计算值。

__device__ unsigned int __vcmpeq4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：a == b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果它们相等，则结果为 ff，否则为 00。例如，__vcmpeq4(0x1234aba5, 0x1234aba6) 返回 0xffffff00。

返回值

如果 a = b，则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpges2(unsigned int a, unsigned int b)

执行每半字有符号比较：a >= b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果 'a' 部分 >= 'b' 部分（有符号比较），则结果为 ffff，否则为 0000。例如，__vcmpges2(0x1234aba5, 0x1234aba6) 返回 0xffff0000。

返回值

如果 a >= b，则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpges4(unsigned int a, unsigned int b)

执行每字节有符号比较：a >= b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果 'a' 部分 >= 'b' 部分（有符号比较），则结果为 ff，否则为 00。例如，__vcmpges4(0x1234aba5, 0x1234aba6) 返回 0xffffff00。

返回值

如果 a >= b，则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpgeu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a >= b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果 'a' 部分 >= 'b' 部分（无符号比较），则结果为 ffff，否则为 0000。例如，__vcmpgeu2(0x1234aba5, 0x1234aba6) 返回 0xffff0000。

返回值

如果 a >= b，则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpgeu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a >= b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果 'a' 部分 >= 'b' 部分（无符号比较），则结果为 ff，否则为 00。例如，__vcmpgeu4(0x1234aba5, 0x1234aba6) 返回 0xffffff00。

返回值

如果 a >= b，则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpgts2(unsigned int a, unsigned int b)

执行每半字有符号比较：a > b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果 'a' 部分 > 'b' 部分（有符号比较），则结果为 ffff，否则为 0000。例如，__vcmpgts2(0x1234aba5, 0x1234aba6) 返回 0x00000000。(注意：示例结果似乎与描述不一致)。

返回值

返回值：如果 a > b（有符号比较），则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpgts4(unsigned int a, unsigned int b)

执行每字节有符号比较：a > b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果 'a' 部分 > 'b' 部分（有符号比较），则结果为 ff，否则为 00。例如，__vcmpgts4(0x1234aba5, 0x1234aba6) 返回 0x00000000。(注意：示例结果似乎与描述不一致)。

返回值

返回值：如果 a > b（有符号比较），则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpgtu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a > b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果 'a' 部分 > 'b' 部分（无符号比较），则结果为 ffff，否则为 0000。例如，__vcmpgtu2(0x1234aba5, 0x1234aba6) 返回 0x00000000。(注意：示例结果似乎与描述不一致)。

返回值

返回值：如果 a > b（有符号比较），则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpgtu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a > b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果 'a' 部分 > 'b' 部分（无符号比较），则结果为 ff，否则为 00。例如，__vcmpgtu4(0x1234aba5, 0x1234aba6) 返回 0x00000000。(注意：示例结果似乎与描述不一致)。

返回值

返回值：如果 a > b（有符号比较），则返回 0xff，否则返回 0。

__device__ unsigned int __vcmples2(unsigned int a, unsigned int b)

执行每半字有符号比较：a <= b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果 'a' 部分 <= 'b' 部分（有符号比较），则结果为 ffff，否则为 0000。例如，__vcmples2(0x1234aba5, 0x1234aba6) 返回 0xffffffff。

返回值

返回值：如果 a <= b（有符号比较），则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmples4(unsigned int a, unsigned int b)

执行每字节有符号比较：a <= b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果 'a' 部分 <= 'b' 部分（有符号比较），则结果为 ff，否则为 00。例如，__vcmples4(0x1234aba5, 0x1234aba6) 返回 0xffffffff。

返回值

返回值：如果 a <= b（有符号比较），则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpleu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a <= b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每个部分包含 2 个字节。对于对应的部分，如果 'a' 部分 <= 'b' 部分（无符号比较），则结果为 ffff，否则为 0000。例如，__vcmpleu2(0x1234aba5, 0x1234aba6) 返回 0xffffffff。

返回值

返回值：如果 a <= b（有符号比较），则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpleu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a <= b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每个部分包含 1 个字节。对于对应的部分，如果 'a' 部分 <= 'b' 部分（无符号比较），则结果为 ff，否则为 00。例如，__vcmpleu4(0x1234aba5, 0x1234aba6) 返回 0xffffffff。

返回值

返回值：如果 a <= b（有符号比较），则返回 0xff，否则返回 0。

__device__ unsigned int __vcmplts2(unsigned int a, unsigned int b)

执行每半字有符号比较：a < b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。对于对应的部分，如果 'a' 部分 < 'b' 部分，则结果为 ffff，否则为 0000。例如，__vcmplts2(0x1234aba5, 0x1234aba6) 返回 0x0000ffff。

返回值

如果 a < b，则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmplts4(unsigned int a, unsigned int b)

执行每字节有符号比较：a < b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每部分包含 1 个字节。对于对应的部分，如果 'a' 部分 < 'b' 部分，则结果为 ff，否则为 00。例如，__vcmplts4(0x1234aba5, 0x1234aba6) 返回 0x000000ff。

返回值

如果 a < b，则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpltu2(unsigned int a, unsigned int b)

执行每半字无符号比较：a < b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。对于对应的部分，如果 'a' 部分 < 'b' 部分，则结果为 ffff，否则为 0000。例如，__vcmpltu2(0x1234aba5, 0x1234aba6) 返回 0x0000ffff。

返回值

如果 a < b，则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpltu4(unsigned int a, unsigned int b)

执行每字节无符号比较：a < b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每部分包含 1 个字节。对于对应的部分，如果 'a' 部分 < 'b' 部分，则结果为 ff，否则为 00。例如，__vcmpltu4(0x1234aba5, 0x1234aba6) 返回 0x000000ff。

返回值

如果 a < b，则返回 0xff，否则返回 0。

__device__ unsigned int __vcmpne2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：a != b ? 0xffff : 0。

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。对于对应的部分，如果 'a' 部分 != 'b' 部分，则结果为 ffff，否则为 0000。例如，__vcmplts2(0x1234aba5, 0x1234aba6) 返回 0x0000ffff。

返回值

如果 a != b，则返回 0xffff，否则返回 0。

__device__ unsigned int __vcmpne4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：a != b ? 0xff : 0。

将每个参数的 4 个字节拆分为 4 个部分，每部分包含 1 个字节。对于对应的部分，如果 'a' 部分 != 'b' 部分，则结果为 ff，否则为 00。例如，__vcmplts4(0x1234aba5, 0x1234aba6) 返回 0x000000ff。

返回值

如果 a != b，则返回 0xff，否则返回 0。

__device__ unsigned int __vhaddu2(unsigned int a, unsigned int b)

执行每半字无符号平均值计算。

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节，然后计算对应部分的无符号平均值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vhaddu4(unsigned int a, unsigned int b)

计算每字节无符号平均值。

将每个参数的 4 个字节拆分为 4 个部分，每部分包含 1 个字节。然后计算对应部分的无符号平均值。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmax_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(a + b, c)

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为有符号短整数。对于对应的部分，函数执行加法和比较：max(a_part + b_part, c_part)。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmax_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(a + b, c), 0)

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为有符号短整数。对于对应的部分，函数执行加法，然后与 relu 取最大值：max(max(a_part + b_part), c_part), 0)。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ int __viaddmax_s32(const int a, const int b, const int c)

计算 max(a + b, c)

计算有符号整数 a 和 b 的和，并与 c 取最大值。

返回值

返回计算后的值。

__host__ __device__ int __viaddmax_s32_relu(const int a, const int b, const int c)

计算 max(max(a + b, c), 0)

计算有符号整数 a 和 b 的和，并与 c 取最大值。如果结果小于 0，则返回 0。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmax_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(a + b, c)

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为无符号短整数。对于对应的部分，函数执行加法和比较：max(a_part + b_part, c_part)。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmax_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 max(a + b, c)

计算无符号整数 a 和 b 的和，并与 c 取最大值。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmin_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(a + b, c)

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为有符号短整数。对于对应的部分，函数执行加法和比较：min(a_part + b_part, c_part)。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmin_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(min(a + b, c), 0)

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为有符号短整数。对于对应的部分，函数执行加法，然后与 relu 取最小值：max(min(a_part + b_part), c_part), 0)。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ int __viaddmin_s32(const int a, const int b, const int c)

计算 min(a + b, c)

计算有符号整数 a 和 b 的和，并与 c 取最小值。

返回值

返回计算后的值。

__host__ __device__ int __viaddmin_s32_relu(const int a, const int b, const int c)

计算 max(min(a + b, c), 0)

计算有符号整数 a 和 b 的和，并与 c 取最小值。如果结果小于 0，则返回 0。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmin_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(a + b, c)

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为无符号短整数。对于对应的部分，函数执行加法和比较：min(a_part + b_part, c_part)。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __viaddmin_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 min(a + b, c)

计算无符号整数 a 和 b 的和，并与 c 取最小值。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vibmax_s16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 max(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a >= b) 的结果。

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为有符号短整数。对于对应的部分，函数执行最大值运算 ( = max(a_part, b_part) )。部分结果被重新组合并作为无符号整数返回。将 `pred_hi` 指向的值设置为 (a_high_part >= b_high_part) 的值。将 `pred_lo` 指向的值设置为 (a_low_part >= b_low_part) 的值。

返回值

返回计算的值。

__host__ __device__ int __vibmax_s32(const int a, const int b, bool *const pred)

计算 max(a, b)，同时将 pred 指向的值设置为 (a >= b)。

计算两个有符号整数 a 和 b 的最大值。还将 `pred` 指向的值设置为 (a >= b) 的值。

返回值

返回计算的值。

__host__ __device__ unsigned int __vibmax_u16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 max(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a >= b) 的结果。

将每个参数的 4 个字节拆分为 2 个部分，每部分包含 2 个字节。这 2 个字节的部分被解释为无符号短整数。对于对应的部分，函数执行最大值运算 ( = max(a_part, b_part) )。部分结果被重新组合并作为无符号整数返回。将 `pred_hi` 指向的值设置为 (a_high_part >= b_high_part) 的值。将 `pred_lo` 指向的值设置为 (a_low_part >= b_low_part) 的值。

返回值

返回计算的值。

__host__ __device__ unsigned int __vibmax_u32(const unsigned int a, const unsigned int b, bool *const pred)

计算 max(a, b)，同时将 pred 指向的值设置为 (a >= b)。

计算两个无符号整数 a 和 b 的最大值。 并且将 pred 指向的值设置为 (a >= b) 的值。

返回值

返回计算的值。

__host__ __device__ unsigned int __vibmin_s16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 min(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a <= b) 的结果。

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行最大值运算（ = max(a_part, b_part) ）。 部分结果被重新组合并作为无符号整数返回。 将 pred_hi 指向的值设置为 (a_high_part <= b_high_part) 的值。 将 pred_lo 指向的值设置为 (a_low_part <= b_low_part) 的值。

返回值

返回计算的值。

__host__ __device__ int __vibmin_s32(const int a, const int b, bool *const pred)

计算 min(a, b)，同时将 pred 指向的值设置为 (a <= b)。

计算两个有符号整数 a 和 b 的最小值。 并且将 pred 指向的值设置为 (a <= b) 的值。

返回值

返回计算的值。

__host__ __device__ unsigned int __vibmin_u16x2(const unsigned int a, const unsigned int b, bool *const pred_hi, bool *const pred_lo)

执行每半字 min(a, b)，同时将 pred_hi 和 pred_lo 指向的值设置为每半字 (a <= b) 的结果。

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为无符号短整数。 对于对应的部分，函数执行最大值运算（ = max(a_part, b_part) ）。 部分结果被重新组合并作为无符号整数返回。 将 pred_hi 指向的值设置为 (a_high_part <= b_high_part) 的值。 将 pred_lo 指向的值设置为 (a_low_part <= b_low_part) 的值。

返回值

返回计算的值。

__host__ __device__ unsigned int __vibmin_u32(const unsigned int a, const unsigned int b, bool *const pred)

计算 min(a, b)，同时将 pred 指向的值设置为 (a <= b)。

计算两个无符号整数 a 和 b 的最小值。 并且将 pred 指向的值设置为 (a <= b) 的值。

返回值

返回计算的值。

__host__ __device__ unsigned int __vimax3_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(a, b), c)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行三向最大值运算（ = max(max(a_part, b_part), c_part) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimax3_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(max(a, b), c), 0)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行带 ReLU 的三向最大值运算（ = max(a_part, b_part, c_part, 0) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ int __vimax3_s32(const int a, const int b, const int c)

计算 max(max(a, b), c)

计算有符号整数 a、b 和 c 的三向最大值。

返回值

返回计算后的值。

__host__ __device__ int __vimax3_s32_relu(const int a, const int b, const int c)

计算 max(max(max(a, b), c), 0)

计算三个有符号整数的最大值，如果该最大值小于 0，则返回 0。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimax3_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(max(a, b), c)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为无符号短整数。 对于对应的部分，函数执行三向最大值运算（ = max(max(a_part, b_part), c_part) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimax3_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 max(max(a, b), c)

计算无符号整数 a、b 和 c 的三向最大值。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimax_s16x2_relu(const unsigned int a, const unsigned int b)

执行每半字 max(max(a, b), 0)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行带 ReLU 的最大值运算（ = max(a_part, b_part, 0) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ int __vimax_s32_relu(const int a, const int b)

计算 max(max(a, b), 0)

计算两个有符号整数 a 和 b 的最大值，如果该最大值小于 0，则返回 0。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimin3_s16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(min(a, b), c)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行三向最小值运算（ = min(min(a_part, b_part), c_part) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimin3_s16x2_relu(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 max(min(min(a, b), c), 0)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行带 ReLU 的三向最小值运算（ = max(min(a_part, b_part, c_part), 0) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ int __vimin3_s32(const int a, const int b, const int c)

计算 min(min(a, b), c)

计算有符号整数 a、b 和 c 的三向最小值。

返回值

返回计算后的值。

__host__ __device__ int __vimin3_s32_relu(const int a, const int b, const int c)

计算 max(min(min(a, b), c), 0)

计算三个有符号整数的最小值，如果该最小值小于 0，则返回 0。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimin3_u16x2(const unsigned int a, const unsigned int b, const unsigned int c)

执行每半字 min(min(a, b), c)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为无符号短整数。 对于对应的部分，函数执行三向最小值运算（ = min(min(a_part, b_part), c_part) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimin3_u32(const unsigned int a, const unsigned int b, const unsigned int c)

计算 min(min(a, b), c)

计算无符号整数 a、b 和 c 的三向最小值。

返回值

返回计算后的值。

__host__ __device__ unsigned int __vimin_s16x2_relu(const unsigned int a, const unsigned int b)

执行每半字 max(min(a, b), 0)

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 这两个字节的部分被解释为有符号短整数。 对于对应的部分，函数执行带 ReLU 的最小值运算（ = max(min(a_part, b_part), 0) ）。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__host__ __device__ int __vimin_s32_relu(const int a, const int b)

计算 max(min(a, b), 0)

计算两个有符号整数 a 和 b 的最小值，如果该最小值小于 0，则返回 0。

返回值

返回计算后的值。

__device__ unsigned int __vmaxs2(unsigned int a, unsigned int b)

执行每半字有符号最大值计算。

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 对于对应的部分，函数计算有符号最大值。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vmaxs4(unsigned int a, unsigned int b)

计算每字节有符号最大值。

将每个参数的 4 个字节拆分为 4 个部分，每个部分由 1 个字节组成。 对于对应的部分，函数计算有符号最大值。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vmaxu2(unsigned int a, unsigned int b)

执行每半字无符号最大值计算。

将每个参数的 4 个字节拆分为 2 个部分，每个部分由 2 个字节组成。 对于对应的部分，函数计算无符号最大值。 部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vmaxu4(unsigned int a, unsigned int b)

计算每字节无符号最大值。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数计算无符号最大值。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vmins2(unsigned int a, unsigned int b)

执行每半字有符号最小值计算。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数计算有符号最小值。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vmins4(unsigned int a, unsigned int b)

计算每字节有符号最小值。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数计算有符号最小值。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vminu2(unsigned int a, unsigned int b)

执行每半字无符号最小值计算。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数计算无符号最小值。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vminu4(unsigned int a, unsigned int b)

计算每字节无符号最小值。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数计算无符号最小值。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vneg2(unsigned int a)

计算每半字取反。

将参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于每个部分，函数计算取反。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vneg4(unsigned int a)

执行每字节取反。

将参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于每个部分，函数计算取反。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vnegss2(unsigned int a)

计算每半字带符号饱和的取反。

将参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于每个部分，函数计算取反。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vnegss4(unsigned int a)

执行每字节带符号饱和的取反。

将参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于每个部分，函数计算取反。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsads2(unsigned int a, unsigned int b)

执行每半字有符号绝对差值之和。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数计算绝对差值并求和。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsads4(unsigned int a, unsigned int b)

计算每字节有符号绝对差值之和。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数计算绝对差值并求和。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsadu2(unsigned int a, unsigned int b)

计算每半字无符号绝对差值之和。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数计算绝对差值，并返回这些差值的和。

返回值

返回计算后的值。

__device__ unsigned int __vsadu4(unsigned int a, unsigned int b)

计算每字节无符号绝对差值之和。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数计算绝对差值，并返回这些差值的和。

返回值

返回计算后的值。

__device__ unsigned int __vseteq2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：如果两个部分都相等，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 == ‘b’ 部分的比较。如果两个相等性都成立，函数返回 1。

返回值

如果 a 等于 b，则返回 1，否则返回 0。

__device__ unsigned int __vseteq4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：如果所有 4 对都相等，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 == ‘b’ 部分的比较。如果两个相等性都成立，函数返回 1。

返回值

如果 a 等于 b，则返回 1，否则返回 0。

__device__ unsigned int __vsetges2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都大于或等于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 >= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a >= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetges4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都大于或等于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 >= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a >= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetgeu2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都大于或等于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 >= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a >= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetgeu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都大于或等于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 >= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a >= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetgts2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都大于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 > ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a > b，则返回 1，否则返回 0。

__device__ unsigned int __vsetgts4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都大于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 > ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a > b，则返回 1，否则返回 0。

__device__ unsigned int __vsetgtu2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都大于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 > ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a > b，则返回 1，否则返回 0。

__device__ unsigned int __vsetgtu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都大于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 > ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a > b，则返回 1，否则返回 0。

__device__ unsigned int __vsetles2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都小于或等于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a <= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetles4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都小于或等于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a <= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetleu2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都小于或等于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a <= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetleu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都小于或等于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a <= b，则返回 1，否则返回 0。

__device__ unsigned int __vsetlts2(unsigned int a, unsigned int b)

执行每半字有符号比较：如果两个部分都小于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a < b，则返回 1，否则返回 0。

__device__ unsigned int __vsetlts4(unsigned int a, unsigned int b)

执行每字节有符号比较：如果所有 4 对都小于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a < b，则返回 1，否则返回 0。

__device__ unsigned int __vsetltu2(unsigned int a, unsigned int b)

执行每半字无符号比较：如果两个部分都小于，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a < b，则返回 1，否则返回 0。

__device__ unsigned int __vsetltu4(unsigned int a, unsigned int b)

执行每字节无符号比较：如果所有 4 对都小于，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 <= ‘b’ 部分的比较。如果两个不等式都成立，函数返回 1。

返回值

如果 a < b，则返回 1，否则返回 0。

__device__ unsigned int __vsetne2(unsigned int a, unsigned int b)

执行每半字（无）符号比较：如果两个部分都不相等，则返回 1。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行 ‘a’ 部分 != ‘b’ 部分的比较。如果两个条件都成立，函数返回 1。

返回值

如果 a 不等于 b，则返回 1，否则返回 0。

__device__ unsigned int __vsetne4(unsigned int a, unsigned int b)

执行每字节（无）符号比较：如果所有 4 对都不相等，则返回 1。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行 ‘a’ 部分 != ‘b’ 部分的比较。如果两个条件都成立，函数返回 1。

返回值

如果 a 不等于 b，则返回 1，否则返回 0。

__device__ unsigned int __vsub2(unsigned int a, unsigned int b)

执行每半字（无）符号减法，带环绕：a - b。

将每个参数的 4 个字节拆分成 2 个部分，每部分 2 个字节。对于对应的部分，函数执行减法运算。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsub4(unsigned int a, unsigned int b)

执行每字节减法：a - b。

将每个参数的 4 个字节拆分成 4 个部分，每部分 1 个字节。对于对应的部分，函数执行减法运算。部分结果被重组并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsubss2(unsigned int a, unsigned int b)

执行每半字（无）符号减法，带符号饱和：a - b。

将每个参数的 4 个字节拆分为 2 部分，每部分包含 2 个字节。对于对应的部分，函数执行带符号饱和的减法运算。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsubss4(unsigned int a, unsigned int b)

执行每字节带符号饱和的减法：a - b。

将每个参数的 4 个字节拆分为 4 部分，每部分包含 1 个字节。对于对应的部分，函数执行带符号饱和的减法运算。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsubus2(unsigned int a, unsigned int b)

执行每半字带无符号饱和的减法：a - b。

将每个参数的 4 个字节拆分为 2 部分，每部分包含 2 个字节。对于对应的部分，函数执行无符号饱和的减法运算。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。

__device__ unsigned int __vsubus4(unsigned int a, unsigned int b)

执行每字节带无符号饱和的减法：a - b。

将每个参数的 4 个字节拆分为 4 部分，每部分包含 1 个字节。对于对应的部分，函数执行无符号饱和的减法运算。部分结果被重新组合并作为无符号整数返回。

返回值

返回计算后的值。