昇腾CANN开源竞赛，从参赛到获奖的实战攻略

前言昇腾CANN开源社区每年都会办技术竞赛——算子开发赛、模型优化赛、应用创新赛奖金从几千到几万不等。但很多人一听到竞赛就打退堂鼓觉得那是大佬的游戏。实际上昇腾CANN竞赛的门槛没想象中高。掌握Ascend C基础语法、熟悉一两个算子的开发流程就能参赛。关键是搞清楚竞赛的规则、评审标准和常见坑——这些信息散落在社区各个角落cann-competitions仓库把它汇总了。这篇会用最直白的方式把昇腾CANN竞赛的完整参赛路径拆解清楚。竞赛全景昇腾CANN开源竞赛分三个赛道每个赛道的考察重点和难度不同赛道1算子开发赛考察重点用Ascend C写一个高性能算子在NPU上跑出好成绩典型赛题实现一个自定义激活函数算子如SwiGLU优化一个已有算子的性能如FlashAttention的某个变体实现一个NPU原生的信号处理算子难度⭐⭐⭐⭐需要对达芬奇架构有基本理解奖金1-5万元赛道2模型优化赛考察重点用CANN工具链优化一个模型的推理性能典型赛题用ATB加速LLM推理降低首token延迟用AMCT量化模型在精度损失1%的前提下提升吞吐用GE图优化减少算子数量难度⭐⭐⭐不需要写算子但要懂CANN工具链奖金1-3万元赛道3应用创新赛考察重点基于昇腾NPU做一个创新应用典型赛题用昇腾NPU做实时视频分析安防/交通用昇腾NPU做边缘推理工业检测/农业用昇腾NPU做行业解决方案金融/医疗难度⭐⭐门槛最低创意和完整性最重要奖金0.5-2万元参赛全流程注册 → 选赛道 → 开发 → 提交PR → CI检查 → 评审 → 颁奖第一步注册在昇腾CANN开源社区注册账号签署CLA贡献者许可协议# Fork竞赛仓库到自己的账号gitclone https://atomgit.com/your-username/cann-competitions.gitcdcann-competitions# 创建参赛目录mkdir-psubmissions/year/track/team-name第二步开发以算子开发赛为例用Ascend C实现一个自定义算子。以下是往届获奖作品的核心代码结构// swiglu_kernel.cpp - Ascend C实现的SwiGLU算子// SwiGLU(x) x * sigmoid(1.702 * x)是LLaMA等大模型的激活函数#includekernel_operator.hclassSwiGLUKernel{public:__aicore__SwiGLUKernel(){}__aicore__voidInit(GM_ADDR x,GM_ADDR y,uint32_ttotalLength){// 分配GPU/NPU上的bufferxGm.SetGlobalBuffer((__gm__ half*)x,totalLength);yGm.SetGlobalBuffer((__gm__ half*)y,totalLength);// 把输入从Global Memory搬到Local Memory更快pipe.InitBuffer(inQueueX,1,totalLength*sizeof(half));pipe.InitBuffer(outQueueY,1,totalLength*sizeof(half));this-totalLengthtotalLength;}__aicore__voidProcess(){// 1. 从Global Memory搬数据到Local MemoryCopyIn();// 2. 在Local Memory上做计算SwiGLU x * sigmoid(1.702x)Compute();// 3. 把结果从Local Memory搬回Global MemoryCopyOut();}private:__aicore__voidCopyIn(){// 从Global Memory读输入LocalTensorhalfxLocalinQueueX.AllocTensorhalf();DataCopy(xLocal,xGm,totalLength);inQueueX.EnQue(xLocal);}__aicore__voidCompute(){LocalTensorhalfxLocalinQueueX.DeQuehalf();LocalTensorhalfyLocaloutQueueY.AllocTensorhalf();// SwiGLU核心计算y x * sigmoid(1.702 * x)// 先算 1.702 * xLocalTensorhalfscaledxLocal;// 复用bufferMuls(scaled,xLocal,(half)1.702,totalLength);// 再算 sigmoid(1.702x)Sigmoid(yLocal,scaled,totalLength);// 最后 x * sigmoidMul(yLocal,xLocal,yLocal,totalLength);outQueueY.EnQuehalf(yLocal);inQueueX.FreeTensor(xLocal);}__aicore__voidCopyOut(){LocalTensorhalfyLocaloutQueueY.DeQuehalf();DataCopy(yGm,yLocal,totalLength);outQueueY.FreeTensor(yLocal);}private:TPipe pipe;TQueQuePosition::VECIN,1inQueueX;TQueQuePosition::VECOUT,1outQueueY;GlobalTensorhalfxGm,yGm;uint32_ttotalLength;};// 算子入口函数externC__global__ __aicore__voidswiglu_kernel(GM_ADDR x,GM_ADDR y,GM_ADDR workspace,GM_ADDR tiling){SwiGLUKernel op;op.Init(x,y,1024);// 1024个元素op.Process();}代码讲解这是Ascend C的典型算子结构——Init/Process三段式。核心在Compute()函数先用Muls做标量乘法1.702×x再用Sigmoid做激活最后用Mul做逐元素乘法。整个过程在Local Memory类似GPU的Shared Memory上执行比在Global Memory上快10倍以上。DataCopy负责Global↔Local的数据搬运是达芬奇架构的DMA操作。第三步提交PR# 1. 写好算子代码和测试脚本# 目录结构# submissions/2026/operator/team-abc/# ├── swiglu_kernel.cpp # 算子实现# ├── swiglu_build.sh # 编译脚本# ├── test_swiglu.py # 测试脚本# └── README.md # 说明文档# 2. 提交到自己的Forkgitadd.gitcommit-s-mfeat(operator): add SwiGLU kernel for Ascend C Implement SwiGLU activation function using Ascend C. Performance: 1024 elements in 0.02ms, 3.2x faster than PyTorch. Signed-off-by: Your Name youremail.comgitpush origin main# 3. 在AtomGit上创建Pull Request# 标题格式[竞赛年份-赛道] 团队名 - 算子/项目名称# 示例[2026-Operator] Team ABC - SwiGLU Kernel代码讲解git commit -s的-s参数会自动加上Signed-off-by行这是开源社区的DCODeveloper Certificate of Origin要求声明你有权提交这个代码。commit message遵循Conventional Commits格式type(scope): description。第四步CI检查提交PR后CI机器人会自动做以下检查# CI检查项自动执行1. DCO签名检查 → commit必须有Signed-off-by2. CLA签署检查 → 提交者必须签署贡献者协议3. 代码格式检查 → clang-format / black4. 编译检查 → Ascend C代码能否通过编译5. 测试执行 → 运行test_swiglu.py对比精度和性能任何一项失败CI会在PR评论区打❌需要修复后重新push。第五步评审CI通过后评审委员会按以下维度打分算子开发赛维度权重说明功能正确性30%输出精度vs基准FP32 PyTorch的误差性能30%在NPU上的执行时间代码质量20%可读性、注释、规范创新性10%是否有独特的优化技巧文档完整性10%README是否清晰往届获奖作品速览年份赛道获奖作品核心优化2025算子开发FlashAttention V3 Ascend CTiling优化Double Buffer2025模型优化LLaMA-7B INT4量化AMCT量化KV Cache FP82025应用创新NPU实时交通检测多模型级联ATB加速踩坑实录坑1未签CLAPR被自动关闭现象提交PR后1分钟内被机器人关闭评论“Please sign the CLA first.”原因昇腾CANN社区要求所有贡献者签署CLAContributor License Agreement未签署的PR会被自动拦截。解决先在AtomGit上签署CLA再提交PR。# 签署CLA# 访问 https://atomgit.com/cann/cann-agreements# 点击Sign CLA按钮用AtomGit账号授权# 签署后重新提交gitcommit--amend-s# 确保有Signed-off-bygitpush-forigin main坑2commit message格式不规范CI失败现象CI报错Invalid commit message format。原因commit message必须遵循Conventional Commits格式如feat(scope): description。解决修改commit message。# 错误gitcommit-madd swiglu kernel# 没有type和scope# 正确gitcommit-s-mfeat(operator): add SwiGLU kernel implementation坑3测试脚本没有性能基准评审被扣分现象功能正确但性能分数低因为测试脚本只验证了精度没有对比PyTorch基准性能。原因评审需要看到你的算子比PyTorch快多少没有基准数据无法评分。解决在测试脚本中加入性能对比。# 测试脚本要包含性能对比importtimeimporttorch# PyTorch基准xtorch.randn(1,4096).npu()torch.npu.synchronize()t0time.time()for_inrange(100):yx*torch.sigmoid(1.702*x)# PyTorch SwiGLUtorch.npu.synchronize()pytorch_time(time.time()-t0)/100# Ascend C算子torch.npu.synchronize()t0time.time()for_inrange(100):yswiglu_custom(x)# 自定义算子torch.npu.synchronize()custom_time(time.time()-t0)/100print(fPyTorch:{pytorch_time*1000:.3f}ms)print(fAscend C:{custom_time*1000:.3f}ms)print(f加速比:{pytorch_time/custom_time:.1f}x)结尾cann-competitions是昇腾CANN开源竞赛的管理仓库涵盖算子开发赛、模型优化赛、应用创新赛三个赛道从注册到评审的完整流程都有规范指导。如果想参与昇腾CANN竞赛建议从应用创新赛入门门槛最低再挑战模型优化赛和算子开发赛。关键是签CLA、commit格式规范、测试脚本要有性能对比数据。昇腾CANN的开源生态还在持续壮大。如果在参赛过程中遇到啥问题欢迎去AtomGit上的昇腾CANN开源社区逛逛里面有一手资料和活跃社区。社区链接https://atomgit.com/cann/cann-competitions