CANN/asc-devkit:__cvta_global_to_generic地址转换函数

📅 2026/7/17 12:07:05
CANN/asc-devkit:__cvta_global_to_generic地址转换函数
__cvta_global_to_generic【免费下载链接】asc-devkit本项目是CANN 推出的昇腾AI处理器专用的算子程序开发语言原生支持C和C标准规范主要由类库和语言扩展层构成提供多层级API满足多维场景算子开发诉求。项目地址: 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 训练系列产品不支持功能说明将Global Memory内存空间的地址值转换为对应的指针并返回。函数原型void* __cvta_global_to_generic(size_t rawbits)参数说明表1参数说明参数名输入/输出描述rawbits输入源操作数。返回值说明指向Global Memory内存空间的指针。该接口不校验输入是否为有效Global Memory地址。异常场景示例说明如下输入场景返回值rawbits为0返回指针使用__isGlobal、__isUbuf、__isLocal均判断为0。rawbits为全1返回指针使用__isGlobal判断为1但该指针不是可安全访问的Global Memory地址。约束说明SIMD与SIMT混合编程场景不支持使用该接口。需要包含的头文件使用该接口需要包含simt_api/device_functions.h头文件。#include simt_api/device_functions.h调用示例SIMT编程场景__global__ __launch_bounds__(1024) void kernel__cvta_global_to_generic(uint32_t* dst, uint32_t* src) { int idx threadIdx.x blockIdx.x * blockDim.x; size_t addr __cvta_generic_to_global(src idx); void* ptr __cvta_global_to_generic(addr); dst[idx] *(uint32_t*)ptr; }【免费下载链接】asc-devkit本项目是CANN 推出的昇腾AI处理器专用的算子程序开发语言原生支持C和C标准规范主要由类库和语言扩展层构成提供多层级API满足多维场景算子开发诉求。项目地址: https://gitcode.com/cann/asc-devkit创作声明:本文部分内容由AI辅助生成(AIGC),仅供参考