华为ascend c 示例代码和详细讲解，昇腾 CANN 算子双版本模板代码

下面给逐行详细注释 + 整体逻辑讲解，基于昇腾CANN 双分支（CCE仿真测试 / ACL 设备运行）场景，同时补充昇腾相关概念。

ai生成

// ==============================================
// 宏分支区分编译环境：CCE内核仿真测试 / 标准ACL应用
// ==============================================
// 如果定义了 CCE_KT_TEST：走CCE内核态本地仿真（CPU模拟aicore运行，单元测试用）
#ifdef CCE_KT_TEST
// 引入TIK/CCE仿真库头文件，提供内存分配、内核调用等仿真接口
#include "tikicpulib.h"
// 定义aicore为空宏，适配aicore内核函数修饰符语法
#define aicore
// 未定义 CCE_KT_TEST：走正式ACL应用层，运行在真实昇腾AI芯片上
#else
// 引入昇腾ACL基础接口头文件（设备、流、内存、上下文管理）
#include "acl/acl.h"
// 同样定义aicore空宏，统一两端代码语法
#define aicore
#endif

// ==============================================
// AICORE 内核函数声明（算子/核函数入口）
// ==============================================
// extern "C"：禁用C++名字修饰，保证C/CCE框架能正常调用该函数
// aicore_void：aicore内核侧空返回值类型（由上面空宏兼容）
// HelloWorld：自定义aicore核函数名
// uint8_t* foo：核函数入参，指向设备全局内存(Gm)数据
extern "C" aicore_void HelloWorld(uint8_t* foo);

// 主函数入口，标准程序入口
int32_t main(int32_t argc, char* argv[])
{
    // 定义内存大小：申请 256 字节的全局内存空间
    size_t fooSize = 256;
    // 定义执行块维度：一次启动 8 个aicore执行单元并行计算
    uint32_t blockDim = 8;

    // ==============================================
    // 分支1：CCE_KT_TEST 内核仿真模式（本地CPU调试）
    // ==============================================
#ifdef CCE_KT_TEST
    // 调用CCE仿真库接口，分配芯片全局内存(Gm)，返回内存指针
    // Ascendl::GmAlloc：仿真环境下的全局内存分配接口
    uint8_t* foo = (uint8_t*)Ascendl::GmAlloc(fooSize);

    // ICPU_RUN_KF：CCE仿真框架调用aicore内核函数的宏
    // 参数1：核函数名 HelloWorld
    // 参数2：并行执行单元数 blockDim
    // 参数3：传给核函数的内存指针 foo
    ICPU_RUN_KF(HelloWorld, blockDim, foo);

    // 释放之前分配的全局内存，避免内存泄漏
    Ascendl::GmFree((void *)foo);

    // ==============================================
    // 分支2：标准ACL模式（真实昇腾硬件运行，业务正式环境）
    // ==============================================
#else
    // 1. 初始化ACL框架，整个程序生命周期只调用一次
    // nullptr：使用默认初始化参数
    aclInit(nullptr);

    // 定义ACL上下文句柄：上下文用于管理设备、流、任务等资源
    aclrtContext context;
    // 指定使用的昇腾设备ID，此处使用第0号设备
    int32_t deviceId = 0;
    // 绑定当前线程到指定昇腾设备
    aclrtSetDevice(deviceId);
    // 基于设备ID创建运行上下文
    aclrtCreateContext(&context, deviceId);

    // 定义流句柄：Stream 用于异步调度AI任务、控制任务执行顺序
    aclrtStream stream = nullptr;
    // 创建一个空的计算流
    aclrtCreateStream(&stream);

    // 定义设备侧内存指针，初始置空
    uint8_t* fooDevice = nullptr;

    // 在昇腾设备上分配显存/全局内存
    // &fooDevice：接收内存指针
    // fooSize：内存大小
    // ACL_MEM_MALLOC_HUGE_FIRST：优先使用大页内存，提升访存性能
    aclrtMalloc((void**)&fooDevice, fooSize, ACL_MEM_MALLOC_HUGE_FIRST);

    // 昇腾核函数调用语法 <<<块数, 共享内存大小, 流>>>
    // blockDim：并行执行单元数
    // nullptr：不使用共享内存
    // stream：任务挂载到指定流上异步执行
    // 后面括号：核函数入参（设备内存指针）
    HelloWorld<<<blockDim, nullptr, stream>>>(fooDevice);

    // 流同步：阻塞当前线程，等待流上所有任务执行完毕
    // 保证核函数跑完再执行后续逻辑
    aclrtSynchronizeStream(stream);

    // 逆序释放资源（创建顺序：内存→流→上下文→设备→ACL，释放反过来）
    // 释放设备侧内存
    aclrtFree(fooDevice);
    // 销毁计算流
    aclrtDestroyStream(stream);
    // 销毁运行上下文
    aclrtDestroyContext(context);
    // 解除当前线程与设备的绑定
    aclrtResetDevice(deviceId);
    // 最终销毁ACL框架，程序退出前最后一步
    aclFinalize();
#endif

    // 程序正常退出，返回0
    return 0;
}

一、整体架构总览

这是昇腾CANN 算子双版本模板代码，一套代码兼容两种运行模式：

1. CCE_KT_TEST 模式

   • 用途：本地仿真调试、单元测试
   • 运行环境：PC/服务器 CPU，模拟 AICORE 内核执行
   • 依赖：tikicpulib.h CCE 仿真库
   • 特点：无需真实昇腾芯片，适合算子开发、代码排错

2. 默认 ACL 模式

   • 用途：线上业务、正式部署
   • 运行环境：真实昇腾AI芯片（310P/310B/910 等）
   • 依赖：标准 ACL 运行时库
   • 特点：走硬件加速流程，完整的「上下文-流-内存-任务调度」

二、关键名词通俗解释

1. aicore
   昇腾芯片的AI计算核心，负责执行算子、矩阵运算、并行计算；代码中用空宏 #define aicore 做语法兼容。

2. Gm 全局内存
   昇腾芯片上的设备全局显存，所有aicore核心都能访问，代码中 GmAlloc / aclrtMalloc 都是分配这块内存。

3. blockDim
   并行粒度：blockDim = 8 代表同时启动 8个aicore单元 并行跑同一个核函数。

4. ACL 上下文(context)
   资源容器：管理设备、流、内存、任务队列，一个程序可创建多个上下文。

5. Stream 流
   异步任务队列：算子任务挂载到流上排队执行，支持多任务并行/串行；aclrtSynchronizeStream 等待流内任务全部结束。

6. <<<>>> 语法
   昇腾对标 CUDA 的核函数启动语法，格式：
   核函数<<<块数, 共享内存, 流>>>(入参)

三、ACL 标准执行流程（固定顺序，不能乱）

ACL 模式是昇腾应用开发标准流程，创建 & 释放必须逆序：

1. aclInit 初始化框架
2. aclrtSetDevice 绑定设备
3. aclrtCreateContext 创建上下文
4. aclrtCreateStream 创建流
5. aclrtMalloc 分配设备内存
6. 启动核函数 HelloWorld<<<>>>
7. aclrtSynchronizeStream 等待任务完成
8. 逆序释放：内存 → 流 → 上下文 → 设备 → aclFinalize

四、两大分支使用场景区分

• 开发调试阶段：加编译宏 -DCCE_KT_TEST，用CCE仿真，不用插昇腾卡，快速调试内核逻辑。
• 上线部署阶段：不加宏，走ACL硬件模式，跑在真实昇腾卡上，发挥硬件加速能力。
  ai生成