InitSocState【免费下载链接】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 训练系列产品不支持功能说明头文件路径为basic_api/kernel_operator_common_intf.h。本接口对AI Core的全局状态进行初始化包括AICCube Core与AIVVector Core的公共状态及各自特有状态。由于不同产品的实现存在差异实际执行的初始化项也有所不同。接口涉及的全部初始化操作汇总如下状态类别初始化内容公共状态初始化原子累加状态关闭数据搬运随路原子操作功能。公共状态初始化Mask工作模式为Normal模式。公共状态初始化用于AddDeqRelu/CastDequant/Cast的s322fp16场景的scale量化参数为1。公共状态初始化CTRL寄存器.md#tab-950)中除CTRL[48]以外的比特位为默认值。公共状态初始化原子操作开启位与原子操作类型.md)为无效的原子操作类型。AIC初始化LoadData卷积数据搬运接口调用时Pad填充的数值为0。AIC初始化使用LoadData卷积数据搬运时A1/B1的边界值为0表示无边界。AIV将Mask配置为全1表示所有数都参与计算。AIV重置loop mode的参数。各产品实际执行的初始化项如下表1针对Ascend 950PR/Ascend 950DT本接口执行的初始化项状态类别初始化内容公共状态原子累加状态公共状态Mask工作模式公共状态用于AddDeqRelu/CastDequant/Cast的s322fp16场景的scale量化参数公共状态CTRL寄存器公共状态原子操作开启位与原子操作类型AICLoadData卷积数据搬运接口调用时Pad填充的数值AIVMaskAIVloop mode的参数表2针对Atlas A3 训练系列产品/Atlas A3 推理系列产品本接口执行的初始化项状态类别初始化内容公共状态原子累加状态公共状态Mask工作模式AICLoadData卷积数据搬运接口调用时Pad填充的数值AIC使用LoadData卷积数据搬运时A1/B1的边界值AIVMask表3针对Atlas A2 训练系列产品/Atlas A2 推理系列产品本接口执行的初始化项状态类别初始化内容公共状态原子累加状态公共状态Mask工作模式AICLoadData卷积数据搬运接口调用时Pad填充的数值AIC使用LoadData卷积数据搬运时A1/B1的边界值AIVMask表4针对Atlas 推理系列产品AI Core和Atlas 推理系列产品Vector Core本接口执行的初始化项状态类别初始化内容公共状态原子累加状态函数原型__aicore__ inline void InitSocState()参数说明无返回值说明无约束说明在实际运行中这些值可能被前序执行的算子修改若不调用该接口进行初始化非预期的值可能导致计算结果出现精度错误。例如前序算子使用Counter模式但未重置为Normal模式当前算子以默认的Normal模式设置Mask时会导致Mask设置不符合预期进而引发精度错误。在TPipe框架编程中初始化过程由TPipe完成无需开发者关注在静态Tensor编程的场景中用户必须在Kernel入口处调用此函数来初始化AI Core状态。调用示例__global__ __aicore__ void add_custom(GM_ADDR x, GM_ADDR y, GM_ADDR z) { // 静态Tensor编程方式中需要开发者手动调用InitSocState()接口初始化全局状态寄存器。 AscendC::InitSocState(); KernelAdd op; op.Init(x, y, z); op.Process(); }【免费下载链接】asc-devkit本项目是CANN 推出的昇腾AI处理器专用的算子程序开发语言原生支持C和C标准规范主要由类库和语言扩展层构成提供多层级API满足多维场景算子开发诉求。项目地址: https://gitcode.com/cann/asc-devkit创作声明：本文部分内容由AI辅助生成（AIGC），仅供参考