master

cann-robotFix: Handle 类型由 void* 改为前向声明结构体指针风格，对齐 cuBLAS/ROCm

07a56d5d创建于 2 天前历史提交

文件	最后提交记录	最后更新时间
sasum_host.cpp	Fix: Handle 类型由 void* 改为前向声明结构体指针风格，对齐 cuBLAS/ROCm Co-authored-by: Zitao Wang<wangzitao4@huawei.com> # message auto-generated for no-merge-commit merge: !277 merge issue_110_258 into master Fix: Handle 类型由 void* 改为前向声明结构体指针风格，对齐 cuBLAS/ROCm Created-by: wangzitao_leo Commit-by: Zitao Wang Merged-by: cann-robot Description: ## 描述 将 aclblasHandle_t 和 aclblasLtHandle_t 从 void* 改为前向声明结构体指针（typedef struct _aclblas_handle* aclblasHandle_t），对齐 cuBLAS/ROCm 主流标准，获得编译期类型校验。Descriptor 类型此前已完成结构体指针化改造，本次仅涉及 Handle。清理了所有冗余的 reinterpret_cast 和 trivial 辅助函数。 ### 改动范围 **核心改动（2 个公开头文件）：** - include/cann_ops_blas.h：typedef void* aclblasHandle_t → typedef struct _aclblas_handle* aclblasHandle_t - include/cann_ops_blasLt.h：typedef void* aclblasLtHandle_t → typedef struct aclblasLtHandle* aclblasLtHandle_t **清理改动（90 个文件）：** - 删除 3 个 trivial 辅助函数（ToInternalHandle、ToInternal x2） - 移除 92 处 reinterpret_cast<_aclblas_handle*>(handle) → 直接用 handle - 移除 3 处 reinterpret_cast<aclblasLtHandle*>(lightHandle) → 直接用 lightHandle - 移除 2 处反向 cast（reinterpret_cast<void*>(h) / reinterpret_cast<aclblasLtHandle_t>(h)）→ 直接赋值 - 更新 1 处过时注释 ### ABI 兼容性 指针的二进制表示不变，无 ABI 破坏。 ## 关联的Issue [#110](https://gitcode.com/cann/ops-blas/issues/110) ## 测试 所有 host 侧目标文件（80+ 算子 host、aclblas_auxiliary、blasLt API）零错误编译通过；grep 确认零遗留 reinterpret_cast 和 ToInternalHandle/ToInternal；ABI 无破坏（指针二进制表示不变）。 ## 文档更新 无需文档更新，API 签名不变。 ## 类型标签 - [x] Bug修复 - [ ] 新特性 - [ ] 性能优化 - [ ] 文档更新 - [ ] 其他，请描述： See merge request: cann/ops-blas!277	2 天前
sasum_kernel.cpp	feat(nrm2): 新增 arch35 aclblasSnrm2接口 Co-authored-by: chensi79<chensi79@huawei.com> # message auto-generated for no-merge-commit merge: !181 merge aclblasSnrm2 into master feat(nrm2): 新增 arch35 aclblasSnrm2接口 Created-by: chensi79 Commit-by: chensi79@huawei.com;chensi79 Merged-by: cann-robot Description: ## 描述 为 ascend950（arch35）新增 aclblasSnrm2 接口实现，并配套调整 arch22 命名与测试目录结构。 ### 改动原因 - 现有 aclblasSnrm2 仅有占位声明（include/cann_ops_blas.h），缺少面向 arch35（Ascend 950）的实现 - arch35 的 AIV 多核 + SIMT 编程模型与 arch22 不同，需要独立的双路径 kernel - 历史目录命名 snrm2_* 统一为 nrm2_* ### 改动方法 1. **arch35 aclblasSnrm2 实现** - 接口签名（include/cann_ops_blas.h）：从占位声明 (handle, const int64_t n, uint8_t* x, const int64_t incx, uint8_t* result) 调整为 (handle, int n, const float* x, int incx, float* result) - Host（blas/nrm2/arch35/nrm2_host.cpp）：参数校验、n≤0/incx≤0 早返回 0、Tiling（useCoreNum = min(aivCoreNum, n, 64)）、workspace 复用 aclblasGetEffectiveWorkspace - Kernel（blas/nrm2/arch35/nrm2_kernel.cpp + nrm2_kernel.h）：双路径 + 汇总 - SIMD 路径（incx==1）：snrm2_aiv_kernel 多核 Mul → ReduceSum 后 atomic add 写入 workspace[blockIdx] - SIMT 路径（incx!=1）：snrm2_simt_kernel 多核 asc_vf_call<Snrm2SimtCompute>，warp-style 归约 - 汇总：snrm2_reduce_kernel 单核 ReduceSum + Sqrt - UB 预算：SAFETY_MARGIN=32KB、UB_MAX_CHUNK_FLOATS=27392（基于 BUFFER_NUM=2、248KB UB 推导） 2. **arch22 命名规范化**（纯重命名，内容基本不变） - blas/nrm2/arch22/snrm2_host.cpp → nrm2_host.cpp - blas/nrm2/arch22/snrm2_kernel.cpp → nrm2_kernel.cpp 3. **测试文件**： - 新增 test/nrm2/snrm2/arch35/（snrm2_test.cpp / snrm2_test.csv / snrm2_npu_wrapper.h） - 新增 test/nrm2/snrm2/snrm2_golden.h（cblas_snrm2 参考）、snrm2_param.h 4. **sasum_kernel.cpp（arch35）修正** - maxCopyPadNum 由 (UINT16_MAX+1)/sizeof(float) 改为 UINT16_MAX/sizeof(float)，避免溢出 - SIMT 归约改用 next-power-of-2 算法并补充 (threadIdx.x + s) < blockDim.x 边界检查，支持非 2 的幂 blockDim.x 5. **文档** - 更新 blas/nrm2/README.md（补充 arch22/arch35 双架构、双路径设计、硬件支持矩阵） ## 关联的Issue - https://gitcode.com/cann/ops-blas/issues/190 ## 测试 - 单元测试 test/nrm2/snrm2/arch35/nrm2_test.cpp，CSV 数据驱动，覆盖： - SIMD 路径：n=0/1/8/32/1024/8192（含 32B 对齐与非对齐） - SIMT 路径：incx=2/3/7（含素数 stride） - 错误路径：NullHandle、x/result nullptr、n≤0、incx≤0 - 参考实现 test/nrm2/snrm2/nrm2_golden.h 调用 cblas_snrm2做精度比对 ## 文档更新 - 更新 blas/nrm2/README.md ## 类型标签 - [ ] Bug修复 - [x] 新特性 - [ ] 性能优化 - [ ] 文档更新 - [ ] 其他，请描述： See merge request: cann/ops-blas!181	20 天前
sasum_tiling_data.h	refactor(blas): flatten directory structure and fix README files Co-authored-by: wuyi_huawei<wuyi36@hisilicon.com> # message auto-generated for no-merge-commit merge: !180 merge docs/issue-167-flatten-blas-structure into master refactor(blas): flatten directory structure and fix README files Created-by: zhanghua145 Commit-by: wuyi_huawei Merged-by: cann-robot Description: ## 描述 统一 blas/ 下所有算子的目录结构规范：算子主目录放置 README.md，arch22/arch35 子目录仅存放源码文件。 具体变更： 1. **移除冗余的算子子目录层级**：将 blas/<op>/<variant>/ 扁平化为 blas/<op>/，涉及 35 个算子目录 2. **融合多架构 README**：将 dot、gemv、gemv_batched、scal、swap、trmv 的 s/c 变体 README 合并为统一文档 3. **清理 arch 子目录 README**：删除 dot/arch35、rotm/arch22、sbmv/arch35、spmv/arch22、spmv/arch35、symv/arch35 下的 README.md，将有效内容合并至主目录 README 4. **更新目录结构引用**：修正 29 个算子 README 中的目录结构描述，对齐扁平化后的实际路径 ## 关联的Issue [#167](https://gitcode.com/cann/ops-blas/issues/167) ## 测试 目录结构重构，不涉及功能变更，原有测试用例不受影响。  ## 文档更新 - 融合 6 个算子的多架构 README（dot/gemv/gemv_batched/scal/swap/trmv） - 删除 6 个 arch 子目录 README 并合并内容（dot/rotm/sbmv/spmv/symv） - 更新 29 个算子 README 的目录结构引用 ## 类型标签 - [ ] Bug修复 - [ ] 新特性 - [ ] 性能优化 - [x] 文档更新 - [x] 其他，请描述：目录结构重构 See merge request: cann/ops-blas!180	28 天前