refactor: 统一头文件 include guard 为 #pragma once 并添加 CI lint 规则-ops-blas-AtomGit

M.pre-commit-config.yaml+9-1

	@@ -12,4 +12,12 @@ repos:
12	rev: v16.0.0	12	rev: v16.0.0
13	hooks:	13	hooks:
14	- id: clang-format	14	- id: clang-format
15	- types_or: [c++, c]	15	+ types_or: [c++, c]
		16	+ - repo: local
		17	+ hooks:
		18	+ - id: check-pragma-once
		19	+ name: Check #pragma once in headers
		20	+ entry: scripts/ci/check_pragma_once.sh
		21	+ language: script
		22	+ types_or: [c, c++]
		23	+ files: '\.(h\|hpp)$'

Magent/skills/blas-op-templates/references/simd-membase/blas/{family}/{op}/archxx/op_tiling_data.h+1-3

	@@ -12,8 +12,7 @@
12	// - host 侧计算后通过 GM 传递给 kernel	12	// - host 侧计算后通过 GM 传递给 kernel
13	// - 字段完全由算子的 Tiling 策略决定	13	// - 字段完全由算子的 Tiling 策略决定
14		14
15	-#ifndef {{OP}}_TILING_DATA_H	15	+#pragma once
16	-#define {{OP}}_TILING_DATA_H
17		16
18	#include <cstdint>	17	#include <cstdint>
19		18
	@@ -33,4 +32,3 @@ struct {{Op}}TilingData {
33	// 建议：使用算子实际的数据类型，而非硬编码 float	32	// 建议：使用算子实际的数据类型，而非硬编码 float
34	constexpr uint32_t ELEMENTS_PER_BLOCK = 32 / sizeof(/* dtype */float);	33	constexpr uint32_t ELEMENTS_PER_BLOCK = 32 / sizeof(/* dtype */float);
35		34
36	-#endif // {{OP}}_TILING_DATA_H

Magent/skills/blas-op-templates/references/simd-regbase/blas/{family}/{op}/arch35/op_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef {{OP}}_TILING_DATA_H	11	+#pragma once
12	-#define {{OP}}_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -40,4 +39,3 @@ struct {{Op}}TilingData {
40		39
41	#pragma pack(pop)	40	#pragma pack(pop)
42		41
43	-#endif // {{OP}}_TILING_DATA_H

Mblas/asum/sasum/arch35/sasum_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SASUM_TILING_DATA_H	11	+#pragma once
12	-#define SASUM_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include "common/helper/kernel_constant.h"	14	#include "common/helper/kernel_constant.h"
	@@ -27,4 +26,3 @@ struct SasumTilingData {
27	uint32_t nthreads;	26	uint32_t nthreads;
28	};	27	};
29		28
30	-#endif // SASUM_TILING_DATA_H

Mblas/common/arch/hardware.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_HARDWARE_H	11	+#pragma once
12	-#define INCLUDE_HARDWARE_H
13		12
14	enum class ArchType { ASCEND_V220, ASCEND_V200, ASCEND_M200 };	13	enum class ArchType { ASCEND_V220, ASCEND_V200, ASCEND_M200 };
15		14
	@@ -32,4 +31,3 @@ struct HardwareInfo {
32	static uint32_t const fbBlockSize = 128;	31	static uint32_t const fbBlockSize = 128;
33	};	32	};
34		33
35	-#endif

Mblas/common/arch/layout.h+1-3

	@@ -8,9 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_LAYOUT_H	11	+#pragma once
12	-#define INCLUDE_LAYOUT_H
13		12
14	enum class DataFormat { ND = 0, NZ, ZN, ZZ, NN, VECTOR };	13	enum class DataFormat { ND = 0, NZ, ZN, ZZ, NN, VECTOR };
15		14
16	-#endif

Mblas/common/compute/mma.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_MMA_H	11	+#pragma once
12	-#define INCLUDE_MMA_H
13		12
14	#include "../arch/hardware.h"	13	#include "../arch/hardware.h"
15	#include "kernel_tensor.h"	14	#include "kernel_tensor.h"
	@@ -71,4 +70,3 @@ struct mmad<ArchTag, ElementA, ElementB, AccDTypeC, false> {
71	};	70	};
72	};	71	};
73		72
74	-#endif

Mblas/common/compute/set_fpc.h+2-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_SET_FPC_H	11	+#pragma once
12	-#define INCLUDE_SET_FPC_H
13		12
14	#include "../arch/hardware.h"	13	#include "../arch/hardware.h"
15	#include "kernel_tensor.h"	14	#include "kernel_tensor.h"
	@@ -34,4 +33,4 @@ struct SetQuantPreAddr<ArchType::ASCEND_V220, DataType> {
34	AscendC::SetFixPipeConfigImpl(quantPreTensor);	33	AscendC::SetFixPipeConfigImpl(quantPreTensor);
35	};	34	};
36	};	35	};
37	-#endif	36	+

Mblas/common/compute/simd.h+2-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_SIMD_H	11	+#pragma once
12	-#define INCLUDE_SIMD_H
13		12
14	#include "../arch/hardware.h"	13	#include "../arch/hardware.h"
15	#include "kernel_operator.h"	14	#include "kernel_operator.h"
	@@ -421,4 +420,4 @@ __aicore__ inline void cgadd_v(AscendC::LocalTensor<DType> dst,
421	{	420	{
422	AscendC::BlockReduceSum<DType, false>(dst, src, repeat, 0, dstRepStride, srcBlkStride, srcRepStride);	421	AscendC::BlockReduceSum<DType, false>(dst, src, repeat, 0, dstRepStride, srcBlkStride, srcRepStride);
423	}	422	}
424	-#endif	423	+

Mblas/common/helper/aclblas_logger_manager.h+1-3

	@@ -13,8 +13,7 @@
13	* \brief ops-blas log配置管理模块（不对外暴露）。	13	* \brief ops-blas log配置管理模块（不对外暴露）。
14	*/	14	*/
15		15
16	-#ifndef ACLBLAS_LOGGER_MANAGER_H_	16	+#pragma once
17	-#define ACLBLAS_LOGGER_MANAGER_H_
18		17
19	#include <acl/acl.h>	18	#include <acl/acl.h>
20	#include <cstddef>	19	#include <cstddef>
	@@ -39,4 +38,3 @@ aclblasStatus_t aclblasGetLoggerCallback(aclblasHandle handle, aclblasLogCallbac
39		38
40	}	39	}
41		40
42	-#endif // ACLBLAS_LOGGER_MANAGER_H_

Mblas/common/helper/host_utils.h+1-3

	@@ -13,8 +13,7 @@
13	* \brief Host-side utilities.	13	* \brief Host-side utilities.
14	*/	14	*/
15		15
16	-#ifndef HOST_UTILS_H	16	+#pragma once
17	-#define HOST_UTILS_H
18		17
19	#include <cstdint>	18	#include <cstdint>
20	#include <cstdio>	19	#include <cstdio>
	@@ -66,4 +65,3 @@ static inline R CeilAlign(T1 val, T2 align)
66	return CeilDiv<R>(val, align) * static_cast<R>(align);	65	return CeilDiv<R>(val, align) * static_cast<R>(align);
67	}	66	}
68		67
69	-#endif // HOST_UTILS_H

Mblas/common/helper/kernel_constant.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef KERNEL_CONSTANT_H	11	+#pragma once
12	-#define KERNEL_CONSTANT_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -19,4 +18,3 @@ constexpr uint32_t UB_SIZE = 248 * 1024;
19	constexpr uint32_t SIMT_MIN_THREAD_NUM = 128;	18	constexpr uint32_t SIMT_MIN_THREAD_NUM = 128;
20	constexpr uint32_t SIMT_MAX_THREAD_NUM = 2048;	19	constexpr uint32_t SIMT_MAX_THREAD_NUM = 2048;
21		20
22	-#endif // KERNEL_CONSTANT_H

Mblas/common/helper/kernel_utils.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef KERNEL_UTILS_H	11	+#pragma once
12	-#define KERNEL_UTILS_H
13		12
14	/* ========== Tiling utilities ========== */	13	/* ========== Tiling utilities ========== */
15		14
	@@ -413,4 +412,3 @@ __aicore__ __inline__ void matrix_ubuf2gm(
413		412
414	#endif // KERNEL_UTILS_LITE	413	#endif // KERNEL_UTILS_LITE
415		414
416	-#endif // KERNEL_UTILS_H

Mblas/common/iterator/iterator.h+2-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_ITERTOR_H	11	+#pragma once
12	-#define INCLUDE_ITERTOR_H
13		12
14	#define KERNEL_UTILS_LITE	13	#define KERNEL_UTILS_LITE
15	#include "../helper/kernel_utils.h"	14	#include "../helper/kernel_utils.h"
	@@ -126,4 +125,4 @@ struct l0c_to_l1 {
126	#include "iterators/l1_to_fb_iterator.inc"	125	#include "iterators/l1_to_fb_iterator.inc"
127	#include "iterators/l1_to_l0_iterator.inc"	126	#include "iterators/l1_to_l0_iterator.inc"
128	#include "iterators/l1_to_ub_iterator.inc"	127	#include "iterators/l1_to_ub_iterator.inc"
129	-#endif	128	+

Mblas/common/memory/mem.h+2-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef INCLUDE_MEM_H	11	+#pragma once
12	-#define INCLUDE_MEM_H
13		12
14	#include "../arch/hardware.h"	13	#include "../arch/hardware.h"
15	#include "kernel_event.h"	14	#include "kernel_event.h"
	@@ -79,4 +78,4 @@ public:
79	public:	78	public:
80	AscendC::LocalTensor<uint8_t> tensor[(uint32_t)BufferType::ASCEND_MAX];	79	AscendC::LocalTensor<uint8_t> tensor[(uint32_t)BufferType::ASCEND_MAX];
81	};	80	};
82	-#endif	81	+

Mblas/dot/sdot/arch35/sdot_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SDOT_TILING_DATA_H	11	+#pragma once
12	-#define SDOT_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -24,4 +23,3 @@ struct SdotTilingData {
24	uint32_t calCount[SDOT_MAX_CORE_NUM];	23	uint32_t calCount[SDOT_MAX_CORE_NUM];
25	};	24	};
26		25
27	-#endif // SDOT_TILING_DATA_H

Mblas/gemv/cgemv/arch22/ascblasCgemv_utils.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef __ASCBLASCGEMV_UTILS_H__	11	+#pragma once
12	-#define __ASCBLASCGEMV_UTILS_H__
13		12
14	#ifdef __CCE_KT_TEST__	13	#ifdef __CCE_KT_TEST__
15	#undef __aicore__	14	#undef __aicore__
	@@ -200,4 +199,3 @@ __aicore__ __inline__ __attribute__((always_inline)) void mulSComplex(
200	}	199	}
201	#endif	200	#endif
202		201
203	-#endif

Mblas/gemv_batched/arch35/gemv_batched_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GEMV_BATCHED_TILING_DATA_H	11	+#pragma once
12	-#define GEMV_BATCHED_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -50,4 +49,3 @@ struct GemvBatchedTilingData {
50	};	49	};
51	#pragma pack(pop)	50	#pragma pack(pop)
52		51
53	-#endif // GEMV_BATCHED_TILING_DATA_H

Mblas/gemv_batched/cgemv_batched/arch22/cgemv_batched_kernel_impl.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef CGEMV_BATCHED_KERNEL_IMPL_H	11	+#pragma once
12	-#define CGEMV_BATCHED_KERNEL_IMPL_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include "kernel_operator.h"	14	#include "kernel_operator.h"
	@@ -640,4 +639,3 @@ __aicore__ inline void CgemvBatchedAIV<T, TRANS_T>::Process()
640	}	639	}
641	} // namespace CgemvBatched	640	} // namespace CgemvBatched
642		641
643	-#endif // CGEMV_BATCHED_AIV_H

Mblas/gemv_batched/cgemv_batched/arch22/cgemv_batched_plan.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef CGEMV_BATCHED_PLAN_H	11	+#pragma once
12	-#define CGEMV_BATCHED_PLAN_H
13		12
14	static constexpr uint8_t BUFFER_NUM = 2;	13	static constexpr uint8_t BUFFER_NUM = 2;
15	static constexpr uint32_t TMPBUF_NUM = 3;	14	static constexpr uint32_t TMPBUF_NUM = 3;
	@@ -97,4 +96,3 @@ uint32_t *CreateCgemvBatchedMask(uint32_t m, uint32_t dtype, uint32_t trans)
97	return maskData;	96	return maskData;
98	}	97	}
99		98
100	-#endif

Mblas/geqrf_batched/sgeqrf_batched/arch35/sgeqrf_batched_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef ACLBLAS_SGEQRF_BATCHED_TILING_DATA_H	11	+#pragma once
12	-#define ACLBLAS_SGEQRF_BATCHED_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -26,4 +25,3 @@ struct GeqrfBatchedTilingData {
26	};	25	};
27	#pragma pack(pop)	26	#pragma pack(pop)
28		27
29	-#endif // ACLBLAS_SGEQRF_BATCHED_TILING_DATA_H

Mblas/gerc/cgerc/arch22/cgerc_kernel_impl.h+2-0

	@@ -8,6 +8,8 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
		11	+#pragma once
		12	+
11	#include "kernel_operator.h"	13	#include "kernel_operator.h"
12	#include "common/helper/kernel_utils.h"	14	#include "common/helper/kernel_utils.h"
13	#include "common/iterator/iterator.h"	15	#include "common/iterator/iterator.h"

Mblas/iamax/iamax/arch22/iamax_kernel_impl.h+1-3

	@@ -7,8 +7,7 @@
7	* INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.	7	* INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10	-#ifndef IAMAX_H	10	+#pragma once
11	-#define IAMAX_H
12		11
13	#include <type_traits>	12	#include <type_traits>
14	#include "kernel_operator.h"	13	#include "kernel_operator.h"
	@@ -570,4 +569,3 @@ __aicore__ inline void Iamax<T>::CopyTmpRstOut()
570		569
571	} // namespace IAMAX	570	} // namespace IAMAX
572		571
573	-#endif // IAMAX_H

Mblas/rotm/srotm/arch22/srotm_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SROTM_TILING_DATA_H	11	+#pragma once
12	-#define SROTM_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -29,4 +28,3 @@ struct SrotmTilingData {
29	float h22;	28	float h22;
30	};	29	};
31		30
32	-#endif // SROTM_TILING_DATA_H

Mblas/rotm/srotm/arch35/srotm_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SROTM_TILING_DATA_H	11	+#pragma once
12	-#define SROTM_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -26,4 +25,3 @@ struct SrotmTilingData {
26	float beta2;	25	float beta2;
27	};	26	};
28		27
29	-#endif // SROTM_TILING_DATA_H

Mblas/scal/sscal/arch35/sscal_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSCAL_TILING_DATA_H	11	+#pragma once
12	-#define SSCAL_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include "common/helper/kernel_constant.h"	14	#include "common/helper/kernel_constant.h"
	@@ -34,4 +33,3 @@ struct SscalTilingData {
34	void sscal_kernel_do(uint8_t* x, uint8_t* workSpace, const SscalTilingData& tiling,	33	void sscal_kernel_do(uint8_t* x, uint8_t* workSpace, const SscalTilingData& tiling,
35	uint32_t numBlocks, void *stream);	34	uint32_t numBlocks, void *stream);
36		35
37	-#endif // SSCAL_TILING_DATA_H

Mblas/scalex/arch35/scalex_kernel.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SCALEX_KERNEL_H	11	+#pragma once
12	-#define SCALEX_KERNEL_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -18,4 +17,3 @@
18	void scalex_kernel_do(uint8_t* x, void* alpha, const ScalexTilingData& tiling,	17	void scalex_kernel_do(uint8_t* x, void* alpha, const ScalexTilingData& tiling,
19	uint32_t numBlocks, void* stream);	18	uint32_t numBlocks, void* stream);
20		19
21	-#endif // SCALEX_KERNEL_H

Mblas/scalex/arch35/scalex_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SCALEX_TILING_DATA_H	11	+#pragma once
12	-#define SCALEX_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -31,4 +30,3 @@ struct ScalexTilingData {
31	uint32_t xType; // aclDataType value	30	uint32_t xType; // aclDataType value
32	};	31	};
33		32
34	-#endif // SCALEX_TILING_DATA_H

Mblas/swap/sswap/arch35/sswap_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSWAP_TILING_DATA_H	11	+#pragma once
12	-#define SSWAP_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -23,4 +22,3 @@ struct SswapTilingData {
23	// elements per 32-byte block for FP32 (= 8)	22	// elements per 32-byte block for FP32 (= 8)
24	constexpr uint32_t ELEMENTS_PER_BLOCK = 32 / sizeof(float);	23	constexpr uint32_t ELEMENTS_PER_BLOCK = 32 / sizeof(float);
25		24
26	-#endif // SSWAP_TILING_DATA_H

Mblas/symm/ssymm/arch22/ssymm_kernel_fwd.h+1-3

	@@ -14,8 +14,7 @@
14	* limitations under the License.	14	* limitations under the License.
15	*/	15	*/
16		16
17	-#ifndef BLAS_SYMM_SSYMM_KERNEL_FWD_H_	17	+#pragma once
18	-#define BLAS_SYMM_SSYMM_KERNEL_FWD_H_
19		18
20	#include <cstdint>	19	#include <cstdint>
21	#include "cann_ops_blas_common.h"	20	#include "cann_ops_blas_common.h"
	@@ -40,4 +39,3 @@ void ssymm_left_cube_do(GM_ADDR aSym, GM_ADDR workspace, GM_ADDR b, GM_ADDR scra
40		39
41	void ssymm_left_cube_postprocess_do(GM_ADDR scratch, GM_ADDR c, GM_ADDR configGm, void *stream);	40	void ssymm_left_cube_postprocess_do(GM_ADDR scratch, GM_ADDR c, GM_ADDR configGm, void *stream);
42		41
43	-#endif // BLAS_SYMM_SSYMM_KERNEL_FWD_H_

Mblas/symm/ssymm/ssymm_common_host.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYMM_COMMON_HOST_H	11	+#pragma once
12	-#define SSYMM_COMMON_HOST_H
13		12
14	#include <cstdio>	13	#include <cstdio>
15	#include <cstdlib>	14	#include <cstdlib>
	@@ -146,4 +145,3 @@ inline void DumpSsymmExecutionBackend(aclblasSideMode_t side,
146	static_cast<int>(side), static_cast<int>(uplo), static_cast<int>(backendKind));	145	static_cast<int>(side), static_cast<int>(uplo), static_cast<int>(backendKind));
147	}	146	}
148		147
149	-#endif // SSYMM_COMMON_HOST_H

Mblas/symm/ssymm/ssymm_common_kernel.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYMM_COMMON_KERNEL_H	11	+#pragma once
12	-#define SSYMM_COMMON_KERNEL_H
13		12
14	#include <cstddef>	13	#include <cstddef>
15	#include <vector>	14	#include <vector>
	@@ -523,4 +522,3 @@ inline RightCubeChunkLocalPolicy BuildRightCubeChunkLocalPolicy(RightCubeUplo up
523	return policy;	522	return policy;
524	}	523	}
525		524
526	-#endif // SSYMM_COMMON_KERNEL_H

Mblas/symm/ssymm/ssymm_common_types.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYMM_COMMON_TYPES_H	11	+#pragma once
12	-#define SSYMM_COMMON_TYPES_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -88,4 +87,3 @@ struct SsymmTilingData {
88	uint32_t rowRemainder; // 剩余行数（前 remainder 个核多分配 1 行）	87	uint32_t rowRemainder; // 剩余行数（前 remainder 个核多分配 1 行）
89	};	88	};
90		89
91	-#endif // SSYMM_COMMON_TYPES_H

Mblas/syr/ssyr/arch35/syr_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SYR_TILING_DATA_H	11	+#pragma once
12	-#define SYR_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -23,4 +22,3 @@ struct SyrTilingData {
23	int64_t incx;	22	int64_t incx;
24	};	23	};
25		24
26	-#endif // SYR_TILING_DATA_H

Mblas/tpttr/stpttr/arch35/stpttr_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPTTR_TILING_DATA_H	11	+#pragma once
12	-#define STPTTR_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -20,4 +19,3 @@ struct TpttrTilingData {
20	uint32_t useCoreNum;	19	uint32_t useCoreNum;
21	};	20	};
22		21
23	-#endif // STPTTR_TILING_DATA_H

Mblas/trttp/strttp/arch35/strttp_tiling_data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STRTTP_TILING_DATA_H	11	+#pragma once
12	-#define STRTTP_TILING_DATA_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -20,4 +19,3 @@ struct TrttpTilingData {
20	uint32_t useCoreNum;	19	uint32_t useCoreNum;
21	};	20	};
22		21
23	-#endif // STRTTP_TILING_DATA_H

Ascripts/ci/check_pragma_once.sh+96-0

@@ -0,0 +1,96 @@
	1	+#!/bin/bash
	2	+# ----------------------------------------------------------------------------
	3	+# Copyright (c) 2026 Huawei Technologies Co., Ltd.
	4	+# This program is free software, you can redistribute it and/or modify it under the terms and conditions of
	5	+# CANN Open Software License Agreement Version 2.0 (the "License").
	6	+# Please refer to the License for details. You may not use this file except in compliance with the License.
	7	+# THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,
	8	+# INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.
	9	+# See LICENSE in the root of the software repository for the full text of the License.
	10	+# ----------------------------------------------------------------------------
	11	+# CI lint 脚本：检查 PR 中新增/修改的 .h/.hpp 文件是否使用 #pragma once
	12	+#
	13	+# 用法（与 run_example.sh 一致，接收 PR 文件列表）：
	14	+# bash scripts/ci/check_pragma_once.sh <pr_filelist.txt>
	15	+#
	16	+# pr_filelist.txt 格式：每行一个文件路径（相对仓库根目录），例如：
	17	+# blas/scal/sscal/arch35/sscal_tiling_data.h
	18	+# test/frame/blas_test.h
	19	+# include/cann_ops_blas.h
	20	+#
	21	+# 仅检查 .h/.hpp 文件，其他文件类型自动跳过。
	22	+#
	23	+# 退出码：
	24	+# 0 — 所有头文件通过检查
	25	+# 1 — 存在使用 #ifndef include guard 或缺少 guard 的头文件
	26	+
	27	+set -euo pipefail
	28	+
	29	+ROOT_PATH=$(cd "$(dirname "$0")"/../.. && pwd)
	30	+cd "${ROOT_PATH}"
	31	+
	32	+if [ $# -lt 1 ]; then
	33	+ echo "Usage: $0 <pr_filelist.txt>"
	34	+ exit 1
	35	+fi
	36	+
	37	+pr_filelist=$1
	38	+
	39	+if [ ! -f "$pr_filelist" ]; then
	40	+ echo "Error: File $pr_filelist not found"
	41	+ exit 1
	42	+fi
	43	+
	44	+failed=0
	45	+checked=0
	46	+
	47	+while IFS= read -r filepath; do
	48	+ # 跳过空行
	49	+ [ -z "$filepath" ] && continue
	50	+
	51	+ # 仅检查 .h / .hpp 文件
	52	+ case "$filepath" in
	53	+ .h\|.hpp) ;;
	54	+ *) continue ;;
	55	+ esac
	56	+
	57	+ # 跳过第三方/工具目录
	58	+ case "$filepath" in
	59	+ .opencode/\|asc-devkit/\|cann-recipes-infer/\|tilelang-ascend/\|msprof_trsm\|dev-doc/\|build/\|out/)
	60	+ continue
	61	+ ;;
	62	+ esac
	63	+
	64	+ # 文件不存在（可能是被删除的文件），跳过
	65	+ [ -f "$filepath" ] \|\| continue
	66	+
	67	+ checked=$((checked + 1))
	68	+
	69	+ # 检查是否使用 #pragma once
	70	+ if grep -q '^#pragma once' "$filepath"; then
	71	+ continue
	72	+ fi
	73	+
	74	+ # 使用 #ifndef include guard
	75	+ if grep -q '^#ifndef' "$filepath"; then
	76	+ echo "ERROR: $filepath uses #ifndef include guard. Please replace with #pragma once."
	77	+ failed=1
	78	+ continue
	79	+ fi
	80	+
	81	+ # 完全没有 guard
	82	+ echo "ERROR: $filepath is missing #pragma once. Please add it after the copyright header."
	83	+ failed=1
	84	+done < "$pr_filelist"
	85	+
	86	+echo "[LINT] Checked $checked header file(s) from PR."
	87	+
	88	+if [ $failed -ne 0 ]; then
	89	+ echo ""
	90	+ echo "Include guard check failed. All .h/.hpp files must use '#pragma once'."
	91	+ echo "See Issue #162 for details."
	92	+ exit 1
	93	+fi
	94	+
	95	+echo "[LINT] Include guard check passed."
	96	+exit 0

Mtest/asum/sasum/arch35/sasum_npu_wrapper.h+2-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SASUM_NPU_H	11	+#pragma once
12	-#define SASUM_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -20,6 +19,7 @@
20	#include "cann_ops_blas.h"	19	#include "cann_ops_blas.h"
21	#include "device.h"	20	#include "device.h"
22		21
		22	+// sasum npu wrapper: allocate device buffer and copy host data (nullable)
23	inline std::unique_ptr<DeviceBuffer> tryAllocAndCopy(const void* hostPtr, size_t bytes)	23	inline std::unique_ptr<DeviceBuffer> tryAllocAndCopy(const void* hostPtr, size_t bytes)
24	{	24	{
25	if (hostPtr == nullptr)	25	if (hostPtr == nullptr)
	@@ -64,4 +64,3 @@ inline aclblasStatus_t aclblasSasum_npu(
64	return ret;	64	return ret;
65	}	65	}
66		66
67	-#endif // SASUM_NPU_H

Mtest/asum/sasum/sasum_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SASUM_CPU_H	11	+#pragma once
12	-#define SASUM_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -42,4 +41,3 @@ inline aclblasStatus_t aclblasSasum_cpu(
42	return ACLBLAS_STATUS_SUCCESS;	41	return ACLBLAS_STATUS_SUCCESS;
43	}	42	}
44		43
45	-#endif // SASUM_CPU_H

Mtest/asum/sasum/sasum_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SASUM_PARAM_H	11	+#pragma once
12	-#define SASUM_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -32,4 +31,3 @@ struct SasumParam : public BlasTestParamBase {
32	}	31	}
33	};	32	};
34		33
35	-#endif // SASUM_PARAM_H

Mtest/blasLtMatmul/arch35/blasLtMatmul_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef LTMATMUL_NPU_H	11	+#pragma once
12	-#define LTMATMUL_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -326,4 +325,3 @@ inline aclblasStatus_t aclblasLtMatmul_npu(
326	return ret;	325	return ret;
327	}	326	}
328		327
329	-#endif // LTMATMUL_NPU_H

Mtest/blasLtMatmul/blasLtMatmul_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef LTMATMUL_CPU_H	11	+#pragma once
12	-#define LTMATMUL_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -322,4 +321,3 @@ inline aclblasStatus_t aclblasLtMatmul_cpu(
322	return ACLBLAS_STATUS_SUCCESS;	321	return ACLBLAS_STATUS_SUCCESS;
323	}	322	}
324		323
325	-#endif // LTMATMUL_CPU_H

Mtest/blasLtMatmul/blasLtMatmul_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef LTMATMUL_PARAM_H	11	+#pragma once
12	-#define LTMATMUL_PARAM_H
13		12
14	#include <string>	13	#include <string>
15	#include <algorithm>	14	#include <algorithm>
	@@ -162,4 +161,3 @@ struct LtMatmulParam : public BlasTestParamBase {
162	}	161	}
163	};	162	};
164		163
165	-#endif // LTMATMUL_PARAM_H

Mtest/blasLtMatrixTransform/arch35/blasLtMatrixTransform_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef LT_MATRIX_TRANSFORM_NPU_H	11	+#pragma once
12	-#define LT_MATRIX_TRANSFORM_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -417,4 +416,3 @@ inline aclblasStatus_t aclblasLtMatrixTransform_npu(
417	return ret;	416	return ret;
418	}	417	}
419		418
420	-#endif // LT_MATRIX_TRANSFORM_NPU_H

Mtest/blasLtMatrixTransform/blasLtMatrixTransform_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef LT_MATRIX_TRANSFORM_GOLDEN_H	11	+#pragma once
12	-#define LT_MATRIX_TRANSFORM_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -519,4 +518,3 @@ inline aclblasStatus_t aclblasLtMatrixTransform_cpu(
519	return ACLBLAS_STATUS_SUCCESS;	518	return ACLBLAS_STATUS_SUCCESS;
520	}	519	}
521		520
522	-#endif // LT_MATRIX_TRANSFORM_GOLDEN_H

Mtest/blasLtMatrixTransform/blasLtMatrixTransform_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef LT_MATRIX_TRANSFORM_PARAM_H	11	+#pragma once
12	-#define LT_MATRIX_TRANSFORM_PARAM_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <string>	14	#include <string>
	@@ -189,4 +188,3 @@ struct LtMatrixTransformParam : public BlasTestParamBase {
189	}	188	}
190	};	189	};
191		190
192	-#endif // LT_MATRIX_TRANSFORM_PARAM_H

Mtest/dot/sdot/arch35/sdot_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SDOT_NPU_H	11	+#pragma once
12	-#define SDOT_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -67,4 +66,3 @@ inline aclblasStatus_t aclblasSdot_npu(
67	return ret;	66	return ret;
68	}	67	}
69		68
70	-#endif // SDOT_NPU_H

Mtest/dot/sdot/sdot_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SDOT_CPU_H	11	+#pragma once
12	-#define SDOT_CPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <vector>	14	#include <vector>
	@@ -33,4 +32,3 @@ inline float aclblasSdot_cpu(int n, const float* x, int incx, const float* y, in
33	return static_cast<float>(sum);	32	return static_cast<float>(sum);
34	}	33	}
35		34
36	-#endif // SDOT_CPU_H

Mtest/dot/sdot/sdot_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SDOT_PARAM_H	11	+#pragma once
12	-#define SDOT_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -32,4 +31,3 @@ struct SdotParam : public BlasTestParamBase {
32	}	31	}
33	};	32	};
34		33
35	-#endif // SDOT_PARAM_H

Mtest/frame/blas_test.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef BLAS_TEST_H	11	+#pragma once
12	-#define BLAS_TEST_H
13		12
14	#include <gtest/gtest.h>	13	#include <gtest/gtest.h>
15	#include "acl/acl.h"	14	#include "acl/acl.h"
	@@ -108,4 +107,3 @@ aclblasHandle_t BlasTest<ParamType>::handle_ = nullptr;
108	template <typename ParamType>	107	template <typename ParamType>
109	aclrtStream BlasTest<ParamType>::stream_ = nullptr;	108	aclrtStream BlasTest<ParamType>::stream_ = nullptr;
110		109
111	-#endif // BLAS_TEST_H

Mtest/frame/csv_loader.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef CSV_LOADER_H	11	+#pragma once
12	-#define CSV_LOADER_H
13		12
14	#include <fstream>	13	#include <fstream>
15	#include <iostream>	14	#include <iostream>
	@@ -318,4 +317,3 @@ inline aclblasComputeType_t parseComputeType(const std::string& s)
318	return parseEnum(s, t, ACLBLAS_COMPUTE_32F);	317	return parseEnum(s, t, ACLBLAS_COMPUTE_32F);
319	}	318	}
320		319
321	-#endif // CSV_LOADER_H

Mtest/frame/data.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef DATA_H	11	+#pragma once
12	-#define DATA_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -115,4 +114,3 @@ public:
115	}	114	}
116	};	115	};
117		116
118	-#endif // DATA_H

Mtest/frame/device.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef DEVICE_H	11	+#pragma once
12	-#define DEVICE_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <memory>	14	#include <memory>
	@@ -103,4 +102,3 @@ inline const float* adjustStridedBase(const float* base, int64_t count, int64_t
103	return adjustStridedBase(const_cast<float*>(base), count, stride);	102	return adjustStridedBase(const_cast<float*>(base), count, stride);
104	}	103	}
105		104
106	-#endif // DEVICE_H

Mtest/frame/fill.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef FILL_H	11	+#pragma once
12	-#define FILL_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cfloat>	14	#include <cfloat>
	@@ -590,4 +589,3 @@ inline std::vector<float> makeBlasFp4Random(int64_t size, uint32_t seed = 0)
590	return data;	589	return data;
591	}	590	}
592		591
593	-#endif // FILL_H

Mtest/frame/types.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TYPES_H	11	+#pragma once
12	-#define TYPES_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -31,4 +30,3 @@ struct VerifyConfig {
31	double epsilonForRel = 1e-7;	30	double epsilonForRel = 1e-7;
32	};	31	};
33		32
34	-#endif // TYPES_H

Mtest/frame/verify.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef VERIFY_H	11	+#pragma once
12	-#define VERIFY_H
13		12
14	#include <cmath>	13	#include <cmath>
15	#include <cstdint>	14	#include <cstdint>
	@@ -312,4 +311,3 @@ private:
312	}	311	}
313	};	312	};
314		313
315	-#endif // VERIFY_H

Mtest/gbmv/sgbmv/arch35/sgbmv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GBMV_NPU_H	11	+#pragma once
12	-#define GBMV_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -91,4 +90,3 @@ inline aclblasStatus_t aclblasSgbmv_npu(
91	return ret;	90	return ret;
92	}	91	}
93		92
94	-#endif // GBMV_NPU_H

Mtest/gbmv/sgbmv/sgbmv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GBMV_CPU_H	11	+#pragma once
12	-#define GBMV_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -88,4 +87,3 @@ inline aclblasStatus_t aclblasSgbmv_cpu(
88	return ACLBLAS_STATUS_SUCCESS;	87	return ACLBLAS_STATUS_SUCCESS;
89	}	88	}
90		89
91	-#endif // GBMV_CPU_H

Mtest/gbmv/sgbmv/sgbmv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GBMV_PARAM_H	11	+#pragma once
12	-#define GBMV_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -50,4 +49,3 @@ struct GbmvParam : public BlasTestParamBase {
50	}	49	}
51	};	50	};
52		51
53	-#endif // GBMV_PARAM_H

Mtest/gels_batched/sgels_batched/arch35/sgels_batched_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GELS_BATCHED_NPU_H	11	+#pragma once
12	-#define GELS_BATCHED_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -200,4 +199,3 @@ inline aclblasStatus_t aclblasSgelsBatched_npu_error(
200	return aclblasSgelsBatched(handle, trans, m, n, nrhs, Aarray, lda, Carray, ldc, devInfo, batchSize);	199	return aclblasSgelsBatched(handle, trans, m, n, nrhs, Aarray, lda, Carray, ldc, devInfo, batchSize);
201	}	200	}
202		201
203	-#endif // GELS_BATCHED_NPU_H

Mtest/gels_batched/sgels_batched/sgels_batched_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GELS_BATCHED_GOLDEN_H	11	+#pragma once
12	-#define GELS_BATCHED_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -274,4 +273,3 @@ inline aclblasStatus_t aclblasSgelsBatched_cpu(
274	return ACLBLAS_STATUS_SUCCESS;	273	return ACLBLAS_STATUS_SUCCESS;
275	}	274	}
276		275
277	-#endif // GELS_BATCHED_GOLDEN_H

Mtest/gels_batched/sgels_batched/sgels_batched_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GELS_BATCHED_PARAM_H	11	+#pragma once
12	-#define GELS_BATCHED_PARAM_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <string>	14	#include <string>
	@@ -56,4 +55,3 @@ struct GelsBatchedParam : public BlasTestParamBase {
56	}	55	}
57	};	56	};
58		57
59	-#endif // GELS_BATCHED_PARAM_H

Mtest/gemv/sgemv/arch35/sgemv_npu_wrapper.h+2-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GEMV_SGEMV_NPU_H	11	+#pragma once
12	-#define GEMV_SGEMV_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -20,6 +19,7 @@
20	#include "cann_ops_blas.h"	19	#include "cann_ops_blas.h"
21	#include "device.h"	20	#include "device.h"
22		21
		22	+// sgemv npu wrapper: allocate device buffer and copy host data (nullable)
23	inline std::unique_ptr<DeviceBuffer> tryAllocAndCopy(const void* hostPtr, size_t bytes)	23	inline std::unique_ptr<DeviceBuffer> tryAllocAndCopy(const void* hostPtr, size_t bytes)
24	{	24	{
25	if (hostPtr == nullptr)	25	if (hostPtr == nullptr)
	@@ -67,4 +67,3 @@ inline aclblasStatus_t aclblasSgemv_npu(
67	return ret;	67	return ret;
68	}	68	}
69		69
70	-#endif // GEMV_SGEMV_NPU_H

Mtest/gemv/sgemv/sgemv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GEMV_SGEMV_GOLDEN_H	11	+#pragma once
12	-#define GEMV_SGEMV_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -50,4 +49,3 @@ inline aclblasStatus_t aclblasSgemv_cpu(
50	return ACLBLAS_STATUS_SUCCESS;	49	return ACLBLAS_STATUS_SUCCESS;
51	}	50	}
52		51
53	-#endif // GEMV_SGEMV_GOLDEN_H

Mtest/gemv/sgemv/sgemv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GEMV_SGEMV_PARAM_H	11	+#pragma once
12	-#define GEMV_SGEMV_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -50,4 +49,3 @@ struct SgemvParam : public BlasTestParamBase {
50	}	49	}
51	};	50	};
52		51
53	-#endif // GEMV_SGEMV_PARAM_H

Mtest/gemv_batched/arch35/gemv_batched_npu_wrapper.h+171-173

	@@ -1,173 +1,171 @@
1	-/**	1	+/**
2	- * Copyright (c) 2026 Huawei Technologies Co., Ltd.	2	+ * Copyright (c) 2026 Huawei Technologies Co., Ltd.
3	- * This program is free software, you can redistribute it and/or modify it under the terms and conditions of	3	+ * This program is free software, you can redistribute it and/or modify it under the terms and conditions of
4	- * CANN Open Software License Agreement Version 2.0 (the "License").	4	+ * CANN Open Software License Agreement Version 2.0 (the "License").
5	- * Please refer to the License for details. You may not use this file except in compliance with the License.	5	+ * Please refer to the License for details. You may not use this file except in compliance with the License.
6	- * THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,	6	+ * THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,
7	- * INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.	7	+ * INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.
8	- * See LICENSE in the root of the software repository for the full text of License.	8	+ * See LICENSE in the root of the software repository for the full text of License.
9	- */	9	+ */
10	-	10	+
11	-#ifndef GEMV_BATCHED_NPU_WRAPPER_H	11	+#pragma once
12	-#define GEMV_BATCHED_NPU_WRAPPER_H	12	+
13	-	13	+#include <algorithm>
14	-#include <algorithm>	14	+#include <cmath>
15	-#include <cmath>	15	+#include <cstdint>
16	-#include <cstdint>	16	+
17	-	17	+#include "acl/acl.h"
18	-#include "acl/acl.h"	18	+#include "cann_ops_blas.h"
19	-#include "cann_ops_blas.h"	19	+#include "device.h"
20	-#include "device.h"	20	+
21	-	21	+// ---- dtype 1 (S): float in, float out ----
22	-// ---- dtype 1 (S): float in, float out ----	22	+inline aclblasStatus_t aclblasGemvBatchedS_npu(
23	-inline aclblasStatus_t aclblasGemvBatchedS_npu(	23	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const float* a, int lda,
24	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const float* a, int lda,	24	+ const float* x, int incx, const float* beta, float* y, int incy, int batchCount)
25	- const float* x, int incx, const float* beta, float* y, int incy, int batchCount)	25	+{
26	-{	26	+ if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;
27	- if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;	27	+ if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {
28	- if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {	28	+ return aclblasSgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);
29	- return aclblasSgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);	29	+ }
30	- }	30	+ if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr)
31	- if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr)	31	+ return ACLBLAS_STATUS_INVALID_VALUE;
32	- return ACLBLAS_STATUS_INVALID_VALUE;	32	+ const int allocLda = std::max(lda, 1);
33	- const int allocLda = std::max(lda, 1);	33	+ const int xCount = (trans == ACLBLAS_OP_N) ? n : m;
34	- const int xCount = (trans == ACLBLAS_OP_N) ? n : m;	34	+ const int yCount = (trans == ACLBLAS_OP_N) ? m : n;
35	- const int yCount = (trans == ACLBLAS_OP_N) ? m : n;	35	+ const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(float);
36	- const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(float);	36	+ const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(float);
37	- const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(float);	37	+ const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(float);
38	- const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(float);	38	+
39	-	39	+ DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);
40	- DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);	40	+ dA.copyFromHost(a, aBytes);
41	- dA.copyFromHost(a, aBytes);	41	+ dX.copyFromHost(x, xBytes);
42	- dX.copyFromHost(x, xBytes);	42	+ dY.copyFromHost(y, yBytes);
43	- dY.copyFromHost(y, yBytes);	43	+ aclblasStatus_t ret = aclblasSgemvBatched(
44	- aclblasStatus_t ret = aclblasSgemvBatched(	44	+ handle, trans, m, n, alpha, static_cast<const float*>(dA.ptr()), lda,
45	- handle, trans, m, n, alpha, static_cast<const float*>(dA.ptr()), lda,	45	+ static_cast<const float*>(dX.ptr()), incx, beta,
46	- static_cast<const float*>(dX.ptr()), incx, beta,	46	+ static_cast<float*>(dY.ptr()), incy, batchCount);
47	- static_cast<float*>(dY.ptr()), incy, batchCount);	47	+ aclrtSynchronizeDevice();
48	- aclrtSynchronizeDevice();	48	+ dY.copyToHost(y, yBytes);
49	- dY.copyToHost(y, yBytes);	49	+ return ret;
50	- return ret;	50	+}
51	-}	51	+
52	-	52	+// ---- dtype 0 (HSH): uint16_t in, uint16_t out ----
53	-// ---- dtype 0 (HSH): uint16_t in, uint16_t out ----	53	+inline aclblasStatus_t aclblasGemvBatchedHSH_npu(
54	-inline aclblasStatus_t aclblasGemvBatchedHSH_npu(	54	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,
55	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,	55	+ const uint16_t* x, int incx, const float* beta, uint16_t* y, int incy, int batchCount)
56	- const uint16_t* x, int incx, const float* beta, uint16_t* y, int incy, int batchCount)	56	+{
57	-{	57	+ if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;
58	- if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;	58	+ if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {
59	- if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {	59	+ return aclblasHSHgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);
60	- return aclblasHSHgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);	60	+ }
61	- }	61	+ if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;
62	- if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;	62	+ const int allocLda = std::max(lda, 1);
63	- const int allocLda = std::max(lda, 1);	63	+ const int xCount = (trans == ACLBLAS_OP_N) ? n : m;
64	- const int xCount = (trans == ACLBLAS_OP_N) ? n : m;	64	+ const int yCount = (trans == ACLBLAS_OP_N) ? m : n;
65	- const int yCount = (trans == ACLBLAS_OP_N) ? m : n;	65	+ const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);
66	- const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);	66	+ const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);
67	- const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);	67	+ const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(uint16_t);
68	- const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(uint16_t);	68	+
69	-	69	+ DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);
70	- DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);	70	+ dA.copyFromHost(a, aBytes);
71	- dA.copyFromHost(a, aBytes);	71	+ dX.copyFromHost(x, xBytes);
72	- dX.copyFromHost(x, xBytes);	72	+ dY.copyFromHost(y, yBytes);
73	- dY.copyFromHost(y, yBytes);	73	+ aclblasStatus_t ret = aclblasHSHgemvBatched(
74	- aclblasStatus_t ret = aclblasHSHgemvBatched(	74	+ handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,
75	- handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,	75	+ static_cast<const uint16_t*>(dX.ptr()), incx, beta,
76	- static_cast<const uint16_t*>(dX.ptr()), incx, beta,	76	+ static_cast<uint16_t*>(dY.ptr()), incy, batchCount);
77	- static_cast<uint16_t*>(dY.ptr()), incy, batchCount);	77	+ aclrtSynchronizeDevice();
78	- aclrtSynchronizeDevice();	78	+ dY.copyToHost(y, yBytes);
79	- dY.copyToHost(y, yBytes);	79	+ return ret;
80	- return ret;	80	+}
81	-}	81	+
82	-	82	+// ---- dtype 2 (HSS): uint16_t in, float out ----
83	-// ---- dtype 2 (HSS): uint16_t in, float out ----	83	+inline aclblasStatus_t aclblasGemvBatchedHSS_npu(
84	-inline aclblasStatus_t aclblasGemvBatchedHSS_npu(	84	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,
85	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,	85	+ const uint16_t* x, int incx, const float* beta, float* y, int incy, int batchCount)
86	- const uint16_t* x, int incx, const float* beta, float* y, int incy, int batchCount)	86	+{
87	-{	87	+ if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;
88	- if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;	88	+ if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {
89	- if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {	89	+ return aclblasHSSgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);
90	- return aclblasHSSgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);	90	+ }
91	- }	91	+ if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;
92	- if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;	92	+ const int allocLda = std::max(lda, 1);
93	- const int allocLda = std::max(lda, 1);	93	+ const int xCount = (trans == ACLBLAS_OP_N) ? n : m;
94	- const int xCount = (trans == ACLBLAS_OP_N) ? n : m;	94	+ const int yCount = (trans == ACLBLAS_OP_N) ? m : n;
95	- const int yCount = (trans == ACLBLAS_OP_N) ? m : n;	95	+ const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);
96	- const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);	96	+ const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);
97	- const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);	97	+ const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(float);
98	- const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(float);	98	+
99	-	99	+ DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);
100	- DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);	100	+ dA.copyFromHost(a, aBytes);
101	- dA.copyFromHost(a, aBytes);	101	+ dX.copyFromHost(x, xBytes);
102	- dX.copyFromHost(x, xBytes);	102	+ dY.copyFromHost(y, yBytes);
103	- dY.copyFromHost(y, yBytes);	103	+ aclblasStatus_t ret = aclblasHSSgemvBatched(
104	- aclblasStatus_t ret = aclblasHSSgemvBatched(	104	+ handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,
105	- handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,	105	+ static_cast<const uint16_t*>(dX.ptr()), incx, beta,
106	- static_cast<const uint16_t*>(dX.ptr()), incx, beta,	106	+ static_cast<float*>(dY.ptr()), incy, batchCount);
107	- static_cast<float*>(dY.ptr()), incy, batchCount);	107	+ aclrtSynchronizeDevice();
108	- aclrtSynchronizeDevice();	108	+ dY.copyToHost(y, yBytes);
109	- dY.copyToHost(y, yBytes);	109	+ return ret;
110	- return ret;	110	+}
111	-}	111	+
112	-	112	+// ---- dtype 3 (TST): uint16_t in, uint16_t out (bf16) ----
113	-// ---- dtype 3 (TST): uint16_t in, uint16_t out (bf16) ----	113	+inline aclblasStatus_t aclblasGemvBatchedTST_npu(
114	-inline aclblasStatus_t aclblasGemvBatchedTST_npu(	114	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,
115	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,	115	+ const uint16_t* x, int incx, const float* beta, uint16_t* y, int incy, int batchCount)
116	- const uint16_t* x, int incx, const float* beta, uint16_t* y, int incy, int batchCount)	116	+{
117	-{	117	+ if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;
118	- if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;	118	+ if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {
119	- if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {	119	+ return aclblasTSTgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);
120	- return aclblasTSTgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);	120	+ }
121	- }	121	+ if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;
122	- if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;	122	+ const int allocLda = std::max(lda, 1);
123	- const int allocLda = std::max(lda, 1);	123	+ const int xCount = (trans == ACLBLAS_OP_N) ? n : m;
124	- const int xCount = (trans == ACLBLAS_OP_N) ? n : m;	124	+ const int yCount = (trans == ACLBLAS_OP_N) ? m : n;
125	- const int yCount = (trans == ACLBLAS_OP_N) ? m : n;	125	+ const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);
126	- const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);	126	+ const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);
127	- const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);	127	+ const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(uint16_t);
128	- const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(uint16_t);	128	+
129	-	129	+ DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);
130	- DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);	130	+ dA.copyFromHost(a, aBytes);
131	- dA.copyFromHost(a, aBytes);	131	+ dX.copyFromHost(x, xBytes);
132	- dX.copyFromHost(x, xBytes);	132	+ dY.copyFromHost(y, yBytes);
133	- dY.copyFromHost(y, yBytes);	133	+ aclblasStatus_t ret = aclblasTSTgemvBatched(
134	- aclblasStatus_t ret = aclblasTSTgemvBatched(	134	+ handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,
135	- handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,	135	+ static_cast<const uint16_t*>(dX.ptr()), incx, beta,
136	- static_cast<const uint16_t*>(dX.ptr()), incx, beta,	136	+ static_cast<uint16_t*>(dY.ptr()), incy, batchCount);
137	- static_cast<uint16_t*>(dY.ptr()), incy, batchCount);	137	+ aclrtSynchronizeDevice();
138	- aclrtSynchronizeDevice();	138	+ dY.copyToHost(y, yBytes);
139	- dY.copyToHost(y, yBytes);	139	+ return ret;
140	- return ret;	140	+}
141	-}	141	+
142	-	142	+// ---- dtype 4 (TSS): uint16_t in, float out (bf16) ----
143	-// ---- dtype 4 (TSS): uint16_t in, float out (bf16) ----	143	+inline aclblasStatus_t aclblasGemvBatchedTSS_npu(
144	-inline aclblasStatus_t aclblasGemvBatchedTSS_npu(	144	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,
145	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const uint16_t* a, int lda,	145	+ const uint16_t* x, int incx, const float* beta, float* y, int incy, int batchCount)
146	- const uint16_t* x, int incx, const float* beta, float* y, int incy, int batchCount)	146	+{
147	-{	147	+ if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;
148	- if (handle == nullptr) return ACLBLAS_STATUS_HANDLE_IS_NULLPTR;	148	+ if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {
149	- if (m <= 0 \|\| n <= 0 \|\| batchCount <= 0) {	149	+ return aclblasTSSgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);
150	- return aclblasTSSgemvBatched(handle, trans, m, n, alpha, a, lda, x, incx, beta, y, incy, batchCount);	150	+ }
151	- }	151	+ if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;
152	- if (a == nullptr \|\| x == nullptr \|\| y == nullptr \|\| alpha == nullptr \|\| beta == nullptr) return ACLBLAS_STATUS_INVALID_VALUE;	152	+ const int allocLda = std::max(lda, 1);
153	- const int allocLda = std::max(lda, 1);	153	+ const int xCount = (trans == ACLBLAS_OP_N) ? n : m;
154	- const int xCount = (trans == ACLBLAS_OP_N) ? n : m;	154	+ const int yCount = (trans == ACLBLAS_OP_N) ? m : n;
155	- const int yCount = (trans == ACLBLAS_OP_N) ? m : n;	155	+ const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);
156	- const size_t aBytes = static_cast<size_t>(batchCount) * allocLda * std::max(1, n) * sizeof(uint16_t);	156	+ const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);
157	- const size_t xBytes = static_cast<size_t>(batchCount) * ((xCount - 1) * std::abs(incx) + 1) * sizeof(uint16_t);	157	+ const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(float);
158	- const size_t yBytes = static_cast<size_t>(batchCount) * ((yCount - 1) * std::abs(incy) + 1) * sizeof(float);	158	+
159	-	159	+ DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);
160	- DeviceBuffer dA(aBytes), dX(xBytes), dY(yBytes);	160	+ dA.copyFromHost(a, aBytes);
161	- dA.copyFromHost(a, aBytes);	161	+ dX.copyFromHost(x, xBytes);
162	- dX.copyFromHost(x, xBytes);	162	+ dY.copyFromHost(y, yBytes);
163	- dY.copyFromHost(y, yBytes);	163	+ aclblasStatus_t ret = aclblasTSSgemvBatched(
164	- aclblasStatus_t ret = aclblasTSSgemvBatched(	164	+ handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,
165	- handle, trans, m, n, alpha, static_cast<const uint16_t*>(dA.ptr()), lda,	165	+ static_cast<const uint16_t*>(dX.ptr()), incx, beta,
166	- static_cast<const uint16_t*>(dX.ptr()), incx, beta,	166	+ static_cast<float*>(dY.ptr()), incy, batchCount);
167	- static_cast<float*>(dY.ptr()), incy, batchCount);	167	+ aclrtSynchronizeDevice();
168	- aclrtSynchronizeDevice();	168	+ dY.copyToHost(y, yBytes);
169	- dY.copyToHost(y, yBytes);	169	+ return ret;
170	- return ret;	170	+}
171	-}	171	+
172	-
173	-#endif // GEMV_BATCHED_NPU_WRAPPER_H

Mtest/gemv_batched/gemv_batched_golden.h+102-104

	@@ -1,104 +1,102 @@
1	-/**	1	+/**
2	- * Copyright (c) 2026 Huawei Technologies Co., Ltd.	2	+ * Copyright (c) 2026 Huawei Technologies Co., Ltd.
3	- * This program is free software, you can redistribute it and/or modify it under the terms and conditions of	3	+ * This program is free software, you can redistribute it and/or modify it under the terms and conditions of
4	- * CANN Open Software License Agreement Version 2.0 (the "License").	4	+ * CANN Open Software License Agreement Version 2.0 (the "License").
5	- * Please refer to the License for details. You may not use this file except in compliance with the License.	5	+ * Please refer to the License for details. You may not use this file except in compliance with the License.
6	- * THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,	6	+ * THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,
7	- * INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.	7	+ * INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.
8	- * See LICENSE in the root of the software repository for the full text of the License.	8	+ * See LICENSE in the root of the software repository for the full text of the License.
9	- */	9	+ */
10	-	10	+
11	-#ifndef GEMV_BATCHED_GOLDEN_H	11	+#pragma once
12	-#define GEMV_BATCHED_GOLDEN_H	12	+
13	-	13	+#include <algorithm>
14	-#include <algorithm>	14	+#include <cmath>
15	-#include <cmath>	15	+#include <cstdint>
16	-#include <cstdint>	16	+
17	-	17	+#include "acl/acl.h"
18	-#include "acl/acl.h"	18	+#include "cann_ops_blas.h"
19	-#include "cann_ops_blas.h"	19	+
20	-	20	+inline aclblasStatus_t validateGemvBatchedCpuParams(
21	-inline aclblasStatus_t validateGemvBatchedCpuParams(	21	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, int lda, int incx,
22	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, int lda, int incx,	22	+ const float* beta, int incy, int batchCount)
23	- const float* beta, int incy, int batchCount)	23	+{
24	-{	24	+ if (handle == nullptr)
25	- if (handle == nullptr)	25	+ return ACLBLAS_STATUS_NOT_INITIALIZED;
26	- return ACLBLAS_STATUS_NOT_INITIALIZED;	26	+ if (trans != ACLBLAS_OP_N && trans != ACLBLAS_OP_T && trans != ACLBLAS_OP_C)
27	- if (trans != ACLBLAS_OP_N && trans != ACLBLAS_OP_T && trans != ACLBLAS_OP_C)	27	+ return ACLBLAS_STATUS_INVALID_VALUE;
28	- return ACLBLAS_STATUS_INVALID_VALUE;	28	+ if (m < 0 \|\| n < 0 \|\| lda < std::max(1, m))
29	- if (m < 0 \|\| n < 0 \|\| lda < std::max(1, m))	29	+ return ACLBLAS_STATUS_INVALID_VALUE;
30	- return ACLBLAS_STATUS_INVALID_VALUE;	30	+ if (incx == 0 \|\| incy == 0)
31	- if (incx == 0 \|\| incy == 0)	31	+ return ACLBLAS_STATUS_INVALID_VALUE;
32	- return ACLBLAS_STATUS_INVALID_VALUE;	32	+ if (alpha == nullptr \|\| beta == nullptr)
33	- if (alpha == nullptr \|\| beta == nullptr)	33	+ return ACLBLAS_STATUS_INVALID_VALUE;
34	- return ACLBLAS_STATUS_INVALID_VALUE;	34	+ if (batchCount < 0)
35	- if (batchCount < 0)	35	+ return ACLBLAS_STATUS_INVALID_VALUE;
36	- return ACLBLAS_STATUS_INVALID_VALUE;	36	+ return ACLBLAS_STATUS_SUCCESS;
37	- return ACLBLAS_STATUS_SUCCESS;	37	+}
38	-}	38	+
39	-	39	+// Per-batch GEMV: y = alpha * op(A) * x + beta * y
40	-// Per-batch GEMV: y = alpha * op(A) * x + beta * y	40	+// For real types OP_C (conjugate-transpose) is equivalent to OP_T.
41	-// For real types OP_C (conjugate-transpose) is equivalent to OP_T.	41	+static void gemvBatchedCpuOne(
42	-static void gemvBatchedCpuOne(	42	+ aclblasOperation_t trans, int m, int n, float alpha, const float* a, int lda,
43	- aclblasOperation_t trans, int m, int n, float alpha, const float* a, int lda,	43	+ const float* x, int incx, float beta, float* y, int incy)
44	- const float* x, int incx, float beta, float* y, int incy)	44	+{
45	-{	45	+ const bool isTransN = (trans == ACLBLAS_OP_N);
46	- const bool isTransN = (trans == ACLBLAS_OP_N);	46	+ const int xCount = isTransN ? n : m;
47	- const int xCount = isTransN ? n : m;	47	+ const int yCount = isTransN ? m : n;
48	- const int yCount = isTransN ? m : n;	48	+ const int absIncx = std::abs(incx);
49	- const int absIncx = std::abs(incx);	49	+ const int absIncy = std::abs(incy);
50	- const int absIncy = std::abs(incy);	50	+
51	-	51	+ for (int i = 0; i < yCount; i++) {
52	- for (int i = 0; i < yCount; i++) {	52	+ int yIdx = (incy > 0) ? (i * incy) : ((yCount - 1 - i) * absIncy);
53	- int yIdx = (incy > 0) ? (i * incy) : ((yCount - 1 - i) * absIncy);	53	+ y[yIdx] *= beta;
54	- y[yIdx] *= beta;	54	+ }
55	- }	55	+
56	-	56	+ if (isTransN) {
57	- if (isTransN) {	57	+ for (int i = 0; i < m; i++) {
58	- for (int i = 0; i < m; i++) {	58	+ double sum = 0.0;
59	- double sum = 0.0;	59	+ for (int j = 0; j < n; j++) {
60	- for (int j = 0; j < n; j++) {	60	+ int xIdx = (incx > 0) ? (j * incx) : ((xCount - 1 - j) * absIncx);
61	- int xIdx = (incx > 0) ? (j * incx) : ((xCount - 1 - j) * absIncx);	61	+ sum += static_cast<double>(a[i + static_cast<int64_t>(j) * lda]) * static_cast<double>(x[xIdx]);
62	- sum += static_cast<double>(a[i + static_cast<int64_t>(j) * lda]) * static_cast<double>(x[xIdx]);	62	+ }
63	- }	63	+ int yIdx = (incy > 0) ? (i * incy) : ((yCount - 1 - i) * absIncy);
64	- int yIdx = (incy > 0) ? (i * incy) : ((yCount - 1 - i) * absIncy);	64	+ y[yIdx] = static_cast<float>(static_cast<double>(alpha) * sum + static_cast<double>(y[yIdx]));
65	- y[yIdx] = static_cast<float>(static_cast<double>(alpha) * sum + static_cast<double>(y[yIdx]));	65	+ }
66	- }	66	+ } else {
67	- } else {	67	+ for (int j = 0; j < n; j++) {
68	- for (int j = 0; j < n; j++) {	68	+ double sum = 0.0;
69	- double sum = 0.0;	69	+ for (int i = 0; i < m; i++) {
70	- for (int i = 0; i < m; i++) {	70	+ int xIdx = (incx > 0) ? (i * incx) : ((xCount - 1 - i) * absIncx);
71	- int xIdx = (incx > 0) ? (i * incx) : ((xCount - 1 - i) * absIncx);	71	+ sum += static_cast<double>(a[i + static_cast<int64_t>(j) * lda]) * static_cast<double>(x[xIdx]);
72	- sum += static_cast<double>(a[i + static_cast<int64_t>(j) * lda]) * static_cast<double>(x[xIdx]);	72	+ }
73	- }	73	+ int yIdx = (incy > 0) ? (j * incy) : ((yCount - 1 - j) * absIncy);
74	- int yIdx = (incy > 0) ? (j * incy) : ((yCount - 1 - j) * absIncy);	74	+ y[yIdx] = static_cast<float>(static_cast<double>(alpha) * sum + static_cast<double>(y[yIdx]));
75	- y[yIdx] = static_cast<float>(static_cast<double>(alpha) * sum + static_cast<double>(y[yIdx]));	75	+ }
76	- }	76	+ }
77	- }	77	+}
78	-}	78	+
79	-	79	+// Unified CPU golden: computes in float using caller-quantized inputs
80	-// Unified CPU golden: computes in float using caller-quantized inputs	80	+inline aclblasStatus_t aclblasGemvBatched_cpu(
81	-inline aclblasStatus_t aclblasGemvBatched_cpu(	81	+ aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const float* a, int lda,
82	- aclblasHandle_t handle, aclblasOperation_t trans, int m, int n, const float* alpha, const float* a, int lda,	82	+ const float* x, int incx, const float* beta, float* y, int incy, int batchCount)
83	- const float* x, int incx, const float* beta, float* y, int incy, int batchCount)	83	+{
84	-{	84	+ aclblasStatus_t validRet = validateGemvBatchedCpuParams(handle, trans, m, n, alpha, lda, incx, beta, incy, batchCount);
85	- aclblasStatus_t validRet = validateGemvBatchedCpuParams(handle, trans, m, n, alpha, lda, incx, beta, incy, batchCount);	85	+ if (validRet != ACLBLAS_STATUS_SUCCESS)
86	- if (validRet != ACLBLAS_STATUS_SUCCESS)	86	+ return validRet;
87	- return validRet;	87	+ if (m == 0 \|\| n == 0 \|\| batchCount == 0)
88	- if (m == 0 \|\| n == 0 \|\| batchCount == 0)	88	+ return ACLBLAS_STATUS_SUCCESS;
89	- return ACLBLAS_STATUS_SUCCESS;	89	+
90	-	90	+ const bool isTransN = (trans == ACLBLAS_OP_N);
91	- const bool isTransN = (trans == ACLBLAS_OP_N);	91	+ const int xCount = isTransN ? n : m;
92	- const int xCount = isTransN ? n : m;	92	+ const int yCount = isTransN ? m : n;
93	- const int yCount = isTransN ? m : n;	93	+ const size_t xStride = static_cast<size_t>((xCount - 1) * std::abs(incx) + 1);
94	- const size_t xStride = static_cast<size_t>((xCount - 1) * std::abs(incx) + 1);	94	+ const size_t yStride = static_cast<size_t>((yCount - 1) * std::abs(incy) + 1);
95	- const size_t yStride = static_cast<size_t>((yCount - 1) * std::abs(incy) + 1);	95	+ const size_t aStride = static_cast<size_t>(lda) * n;
96	- const size_t aStride = static_cast<size_t>(lda) * n;	96	+
97	-	97	+ for (int b = 0; b < batchCount; b++) {
98	- for (int b = 0; b < batchCount; b++) {	98	+ gemvBatchedCpuOne(trans, m, n, alpha, a + b aStride, lda, x + b * xStride, incx, beta, y + b yStride, incy);
99	- gemvBatchedCpuOne(trans, m, n, alpha, a + b aStride, lda, x + b * xStride, incx, beta, y + b yStride, incy);	99	+ }
100	- }	100	+ return ACLBLAS_STATUS_SUCCESS;
101	- return ACLBLAS_STATUS_SUCCESS;	101	+}
102	-}	102	+
103	-
104	-#endif // GEMV_BATCHED_GOLDEN_H

Mtest/gemv_batched/gemv_batched_param.h+56-58

	@@ -1,58 +1,56 @@
1	-/**	1	+/**
2	- * Copyright (c) 2026 Huawei Technologies Co., Ltd.	2	+ * Copyright (c) 2026 Huawei Technologies Co., Ltd.
3	- * This program is free software, you can redistribute it and/or modify it under the terms and conditions of	3	+ * This program is free software, you can redistribute it and/or modify it under the terms and conditions of
4	- * CANN Open Software License Agreement Version 2.0 (the "License").	4	+ * CANN Open Software License Agreement Version 2.0 (the "License").
5	- * Please refer to the License for details. You may not use this file except in compliance with the License.	5	+ * Please refer to the License for details. You may not use this file except in compliance with the License.
6	- * THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,	6	+ * THIS SOFTWARE IS PROVIDED ON AN "AS IS" BASIS, WITHOUT WARRANTIES OF ANY KIND, EITHER EXPRESS OR IMPLIED,
7	- * INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.	7	+ * INCLUDING BUT NOT LIMITED TO NON-INFRINGEMENT, MERCHANTABILITY, OR FITNESS FOR A PARTICULAR PURPOSE.
8	- * See LICENSE in the root of the software repository for the full text of the License.	8	+ * See LICENSE in the root of the software repository for the full text of the License.
9	- */	9	+ */
10	-	10	+
11	-#ifndef GEMV_BATCHED_PARAM_H	11	+#pragma once
12	-#define GEMV_BATCHED_PARAM_H	12	+
13	-	13	+#include <string>
14	-#include <string>	14	+
15	-	15	+#include "acl/acl.h"
16	-#include "acl/acl.h"	16	+#include "cann_ops_blas.h"
17	-#include "cann_ops_blas.h"	17	+#include "csv_loader.h"
18	-#include "csv_loader.h"	18	+
19	-	19	+struct GemvBatchedParam : public BlasTestParamBase {
20	-struct GemvBatchedParam : public BlasTestParamBase {	20	+ // 0=HSH(F16→F16) 1=S(F32) 2=HSS(F16→F32) 3=TST(BF16→BF16) 4=TSS(BF16→F32)
21	- // 0=HSH(F16→F16) 1=S(F32) 2=HSS(F16→F32) 3=TST(BF16→BF16) 4=TSS(BF16→F32)	21	+ int dtype = 1;
22	- int dtype = 1;	22	+ aclblasOperation_t trans = ACLBLAS_OP_N;
23	- aclblasOperation_t trans = ACLBLAS_OP_N;	23	+ int m = 0;
24	- int m = 0;	24	+ int n = 0;
25	- int n = 0;	25	+ int batchCount = 1;
26	- int batchCount = 1;	26	+ float alpha = 1.0f;
27	- float alpha = 1.0f;	27	+ BlasFillMode alphaFill = parseFill("RANDOM");
28	- BlasFillMode alphaFill = parseFill("RANDOM");	28	+ BlasFillMode a = parseFill("RANDOM");
29	- BlasFillMode a = parseFill("RANDOM");	29	+ int lda = 0;
30	- int lda = 0;	30	+ BlasFillMode x = parseFill("RANDOM");
31	- BlasFillMode x = parseFill("RANDOM");	31	+ int incx = 1;
32	- int incx = 1;	32	+ float beta = 0.0f;
33	- float beta = 0.0f;	33	+ BlasFillMode betaFill = parseFill("RANDOM");
34	- BlasFillMode betaFill = parseFill("RANDOM");	34	+ BlasFillMode y = parseFill("RANDOM");
35	- BlasFillMode y = parseFill("RANDOM");	35	+ int incy = 1;
36	- int incy = 1;	36	+
37	-	37	+ GemvBatchedParam(const csv_map& map) : BlasTestParamBase(map)
38	- GemvBatchedParam(const csv_map& map) : BlasTestParamBase(map)	38	+ {
39	- {	39	+ dtype = parseInt(ReadMap(map, "dtype", "1"));
40	- dtype = parseInt(ReadMap(map, "dtype", "1"));	40	+ trans = parseOpTrans(ReadMap(map, "trans", "N"));
41	- trans = parseOpTrans(ReadMap(map, "trans", "N"));	41	+ m = parseInt(ReadMap(map, "m", "0"));
42	- m = parseInt(ReadMap(map, "m", "0"));	42	+ n = parseInt(ReadMap(map, "n", "0"));
43	- n = parseInt(ReadMap(map, "n", "0"));	43	+ batchCount = parseInt(ReadMap(map, "batchCount", "1"));
44	- batchCount = parseInt(ReadMap(map, "batchCount", "1"));	44	+ alpha = parseFloat(ReadMap(map, "alpha", "1.0"));
45	- alpha = parseFloat(ReadMap(map, "alpha", "1.0"));	45	+ alphaFill = parseFill(ReadMap(map, "alpha_fill", "RANDOM"));
46	- alphaFill = parseFill(ReadMap(map, "alpha_fill", "RANDOM"));	46	+ a = parseFill(ReadMap(map, "a", "RANDOM"));
47	- a = parseFill(ReadMap(map, "a", "RANDOM"));	47	+ lda = parseInt(ReadMap(map, "lda", std::to_string(std::max(1, m))));
48	- lda = parseInt(ReadMap(map, "lda", std::to_string(std::max(1, m))));	48	+ x = parseFill(ReadMap(map, "x", "RANDOM"));
49	- x = parseFill(ReadMap(map, "x", "RANDOM"));	49	+ incx = parseInt(ReadMap(map, "incx", "1"));
50	- incx = parseInt(ReadMap(map, "incx", "1"));	50	+ beta = parseFloat(ReadMap(map, "beta", "0.0"));
51	- beta = parseFloat(ReadMap(map, "beta", "0.0"));	51	+ betaFill = parseFill(ReadMap(map, "beta_fill", "RANDOM"));
52	- betaFill = parseFill(ReadMap(map, "beta_fill", "RANDOM"));	52	+ y = parseFill(ReadMap(map, "y", "RANDOM"));
53	- y = parseFill(ReadMap(map, "y", "RANDOM"));	53	+ incy = parseInt(ReadMap(map, "incy", "1"));
54	- incy = parseInt(ReadMap(map, "incy", "1"));	54	+ }
55	- }	55	+};
56	-};	56	+
57	-
58	-#endif // GEMV_BATCHED_PARAM_H

Mtest/geqrf_batched/sgeqrf_batched/arch35/sgeqrf_batched_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGEQRF_BATCHED_NPU_H	11	+#pragma once
12	-#define SGEQRF_BATCHED_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -134,4 +133,3 @@ inline aclblasStatus_t aclblasSgeqrfBatched_npu(
134	return (ret != ACLBLAS_STATUS_SUCCESS) ? ret : copyRet;	133	return (ret != ACLBLAS_STATUS_SUCCESS) ? ret : copyRet;
135	}	134	}
136		135
137	-#endif // SGEQRF_BATCHED_NPU_H

Mtest/geqrf_batched/sgeqrf_batched/sgeqrf_batched_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGEQRF_BATCHED_GOLDEN_H	11	+#pragma once
12	-#define SGEQRF_BATCHED_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -77,4 +76,3 @@ inline aclblasStatus_t aclblasSgeqrfBatched_cpu(
77	return ACLBLAS_STATUS_SUCCESS;	76	return ACLBLAS_STATUS_SUCCESS;
78	}	77	}
79		78
80	-#endif // SGEQRF_BATCHED_GOLDEN_H

Mtest/geqrf_batched/sgeqrf_batched/sgeqrf_batched_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGEQRF_BATCHED_PARAM_H	11	+#pragma once
12	-#define SGEQRF_BATCHED_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -38,4 +37,3 @@ struct SgeqrfBatchedParam : public BlasTestParamBase {
38	}	37	}
39	};	38	};
40		39
41	-#endif // SGEQRF_BATCHED_PARAM_H

Mtest/ger/sger/arch35/sger_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGER_NPU_H	11	+#pragma once
12	-#define SGER_NPU_H
13		12
14	#include <climits>	13	#include <climits>
15	#include <cstdint>	14	#include <cstdint>
	@@ -91,4 +90,3 @@ inline aclblasStatus_t aclblasSger_npu(
91	return ret;	90	return ret;
92	}	91	}
93		92
94	-#endif // SGER_NPU_H

Mtest/ger/sger/sger_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGER_GOLDEN_H	11	+#pragma once
12	-#define SGER_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -70,4 +69,3 @@ inline aclblasStatus_t aclblasSger_cpu(
70	return ACLBLAS_STATUS_SUCCESS;	69	return ACLBLAS_STATUS_SUCCESS;
71	}	70	}
72		71
73	-#endif // SGER_GOLDEN_H

Mtest/ger/sger/sger_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGER_PARAM_H	11	+#pragma once
12	-#define SGER_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -44,4 +43,3 @@ struct SgerParam : public BlasTestParamBase {
44	}	43	}
45	};	44	};
46		45
47	-#endif // SGER_PARAM_H

Mtest/getrf_batched/sgetrf_batched/arch35/sgetrf_batched_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGETRF_BATCHED_NPU_H	11	+#pragma once
12	-#define SGETRF_BATCHED_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <cstring>	14	#include <cstring>
	@@ -184,4 +183,3 @@ inline aclblasStatus_t aclblasSgetrfBatched_npu(
184	return ret;	183	return ret;
185	}	184	}
186		185
187	-#endif // SGETRF_BATCHED_NPU_H

Mtest/getrf_batched/sgetrf_batched/sgetrf_batched_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGETRF_BATCHED_GOLDEN_H	11	+#pragma once
12	-#define SGETRF_BATCHED_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -110,4 +109,3 @@ inline aclblasStatus_t aclblasSgetrfBatched_cpu(
110	return ACLBLAS_STATUS_SUCCESS;	109	return ACLBLAS_STATUS_SUCCESS;
111	}	110	}
112		111
113	-#endif // SGETRF_BATCHED_GOLDEN_H

Mtest/getrf_batched/sgetrf_batched/sgetrf_batched_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SGETRF_BATCHED_PARAM_H	11	+#pragma once
12	-#define SGETRF_BATCHED_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -84,4 +83,3 @@ struct SgetrfBatchedParam : public BlasTestParamBase {
84	}	83	}
85	};	84	};
86		85
87	-#endif // SGETRF_BATCHED_PARAM_H

Mtest/getri_batched/arch35/getri_batched_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GETRI_BATCHED_NPU_H	11	+#pragma once
12	-#define GETRI_BATCHED_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <cstring>	14	#include <cstring>
	@@ -242,4 +241,3 @@ inline aclblasStatus_t aclblasSgetriBatched_npu(
242	return ret;	241	return ret;
243	}	242	}
244		243
245	-#endif // GETRI_BATCHED_NPU_H

Mtest/getri_batched/getri_batched_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GETRI_BATCHED_GOLDEN_H	11	+#pragma once
12	-#define GETRI_BATCHED_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -141,4 +140,3 @@ inline void aclblasSgetrfBatched_cpu_for_getri(
141	}	140	}
142	}	141	}
143		142
144	-#endif // GETRI_BATCHED_GOLDEN_H

Mtest/getri_batched/getri_batched_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef GETRI_BATCHED_PARAM_H	11	+#pragma once
12	-#define GETRI_BATCHED_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -98,4 +97,3 @@ struct SgetriBatchedParam : public BlasTestParamBase {
98	}	97	}
99	};	98	};
100		99
101	-#endif // GETRI_BATCHED_PARAM_H

Mtest/rotm/srotm/arch22/srotm_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SROTM_NPU_WRAPPER_H	11	+#pragma once
12	-#define SROTM_NPU_WRAPPER_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <memory>	14	#include <memory>
	@@ -72,4 +71,3 @@ inline aclblasStatus_t aclblasSrotm_npu(
72	return ret;	71	return ret;
73	}	72	}
74		73
75	-#endif // SROTM_NPU_WRAPPER_H

Mtest/rotm/srotm/arch35/srotm_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SROTM_NPU_H	11	+#pragma once
12	-#define SROTM_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -65,4 +64,3 @@ inline aclblasStatus_t aclblasSrotm_npu(
65	return ret;	64	return ret;
66	}	65	}
67		66
68	-#endif // SROTM_NPU_H

Mtest/rotm/srotm/srotm_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SROTM_GOLDEN_H	11	+#pragma once
12	-#define SROTM_GOLDEN_H
13		12
14	#include <array>	13	#include <array>
15	#include <cstdint>	14	#include <cstdint>
	@@ -62,4 +61,3 @@ inline aclblasStatus_t aclblasSrotm_cpu(
62	return ACLBLAS_STATUS_SUCCESS;	61	return ACLBLAS_STATUS_SUCCESS;
63	}	62	}
64		63
65	-#endif // SROTM_GOLDEN_H

Mtest/rotm/srotm/srotm_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SROTM_PARAM_H	11	+#pragma once
12	-#define SROTM_PARAM_H
13		12
14	#include <array>	13	#include <array>
15	#include <cstdint>	14	#include <cstdint>
	@@ -38,4 +37,3 @@ struct SrotmParam : public BlasTestParamBase {
38	}	37	}
39	};	38	};
40		39
41	-#endif // SROTM_PARAM_H

Mtest/scal/sscal/arch35/sscal_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSCAL_NPU_H	11	+#pragma once
12	-#define SSCAL_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <algorithm>	14	#include <algorithm>
	@@ -50,4 +49,3 @@ inline aclblasStatus_t aclblasSscal_npu(
50	return ret;	49	return ret;
51	}	50	}
52		51
53	-#endif // SSCAL_NPU_H

Mtest/scal/sscal/sscal_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSCAL_GOLDEN_H	11	+#pragma once
12	-#define SSCAL_GOLDEN_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <cmath>	14	#include <cmath>
	@@ -39,4 +38,3 @@ inline aclblasStatus_t aclblasSscal_cpu(
39	return ACLBLAS_STATUS_SUCCESS;	38	return ACLBLAS_STATUS_SUCCESS;
40	}	39	}
41		40
42	-#endif // SSCAL_GOLDEN_H

Mtest/scal/sscal/sscal_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSCAL_PARAM_H	11	+#pragma once
12	-#define SSCAL_PARAM_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <string>	14	#include <string>
	@@ -34,4 +33,3 @@ struct SscalParam : public BlasTestParamBase {
34	}	33	}
35	};	34	};
36		35
37	-#endif // SSCAL_PARAM_H

Mtest/scalex/arch35/scalex_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SCALEX_NPU_WRAPPER_H	11	+#pragma once
12	-#define SCALEX_NPU_WRAPPER_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <cstdlib>	14	#include <cstdlib>
	@@ -173,4 +172,3 @@ inline aclblasStatus_t aclblasScalex_npu(
173	return ret;	172	return ret;
174	}	173	}
175		174
176	-#endif // SCALEX_NPU_WRAPPER_H

Mtest/scalex/scalex_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SCALEX_GOLDEN_H	11	+#pragma once
12	-#define SCALEX_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -149,4 +148,3 @@ inline aclblasStatus_t aclblasScalex_cpu(
149	return ACLBLAS_STATUS_SUCCESS;	148	return ACLBLAS_STATUS_SUCCESS;
150	}	149	}
151		150
152	-#endif // SCALEX_GOLDEN_H

Mtest/scalex/scalex_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SCALEX_PARAM_H	11	+#pragma once
12	-#define SCALEX_PARAM_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <string>	14	#include <string>
	@@ -60,4 +59,3 @@ struct ScalexParam : public BlasTestParamBase {
60	}	59	}
61	};	60	};
62		61
63	-#endif // SCALEX_PARAM_H

Mtest/swap/sswap/arch35/sswap_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSWAP_NPU_H	11	+#pragma once
12	-#define SSWAP_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -76,4 +75,3 @@ inline aclblasStatus_t aclblasSswap_npu(aclblasHandle_t handle, int n, float* x,
76	return ret;	75	return ret;
77	}	76	}
78		77
79	-#endif // SSWAP_NPU_H

Mtest/swap/sswap/sswap_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSWAP_GOLDEN_H	11	+#pragma once
12	-#define SSWAP_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -35,4 +34,3 @@ inline aclblasStatus_t aclblasSswap_cpu(aclblasHandle_t handle, int n, float* x,
35	return ACLBLAS_STATUS_SUCCESS;	34	return ACLBLAS_STATUS_SUCCESS;
36	}	35	}
37		36
38	-#endif // SSWAP_GOLDEN_H

Mtest/swap/sswap/sswap_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSWAP_PARAM_H	11	+#pragma once
12	-#define SSWAP_PARAM_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <string>	14	#include <string>
	@@ -35,4 +34,3 @@ struct SswapParam : public BlasTestParamBase {
35	}	34	}
36	};	35	};
37		36
38	-#endif // SSWAP_PARAM_H

Mtest/symm/ssymm/arch22/ssymm_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYMM_NPU_WRAPPER_H	11	+#pragma once
12	-#define SSYMM_NPU_WRAPPER_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -38,4 +37,3 @@ inline aclblasStatus_t aclblasSsymm_npu(
38	return aclblasSsymm(handle, side, uplo, m, n, alpha, A, lda, B, ldb, beta, C, ldc);	37	return aclblasSsymm(handle, side, uplo, m, n, alpha, A, lda, B, ldb, beta, C, ldc);
39	}	38	}
40		39
41	-#endif // SSYMM_NPU_WRAPPER_H

Mtest/symm/ssymm/ssymm_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYMM_GOLDEN_H	11	+#pragma once
12	-#define SSYMM_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -104,4 +103,3 @@ inline aclblasStatus_t aclblasSsymm_cpu(
104	return ACLBLAS_STATUS_SUCCESS;	103	return ACLBLAS_STATUS_SUCCESS;
105	}	104	}
106		105
107	-#endif // SSYMM_GOLDEN_H

Mtest/symm/ssymm/ssymm_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYMM_PARAM_H	11	+#pragma once
12	-#define SSYMM_PARAM_H
13		12
14	#include <string>	13	#include <string>
15	#include "acl/acl.h"	14	#include "acl/acl.h"
	@@ -51,4 +50,3 @@ struct SsymmParam : public BlasTestParamBase {
51	}	50	}
52	};	51	};
53		52
54	-#endif // SSYMM_PARAM_H

Mtest/syr/ssyr/arch35/ssyr_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYR_NPU_H	11	+#pragma once
12	-#define SSYR_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -92,4 +91,3 @@ inline aclblasStatus_t aclblasSsyr_npu(
92	return ret;	91	return ret;
93	}	92	}
94		93
95	-#endif // SSYR_NPU_H

Mtest/syr/ssyr/ssyr_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYR_GOLDEN_H	11	+#pragma once
12	-#define SSYR_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -43,4 +42,3 @@ inline void aclblasSsyr_cpu(aclblasFillMode_t uplo, int n, float alpha, const fl
43	}	42	}
44	}	43	}
45		44
46	-#endif // SSYR_GOLDEN_H

Mtest/syr/ssyr/ssyr_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYR_PARAM_H	11	+#pragma once
12	-#define SSYR_PARAM_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <string>	14	#include <string>
	@@ -39,4 +38,3 @@ struct SsyrParam : public BlasTestParamBase {
39	}	38	}
40	};	39	};
41		40
42	-#endif // SSYR_PARAM_H

Mtest/syr2/ssyr2/arch35/ssyr2_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYR2_NPU_H	11	+#pragma once
12	-#define SSYR2_NPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -105,4 +104,3 @@ inline aclblasStatus_t aclblasSsyr2_npu(
105	return ret;	104	return ret;
106	}	105	}
107		106
108	-#endif // SSYR2_NPU_H

Mtest/syr2/ssyr2/ssyr2_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYR2_GOLDEN_H	11	+#pragma once
12	-#define SSYR2_GOLDEN_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -48,4 +47,3 @@ inline void aclblasSsyr2_cpu(
48	}	47	}
49	}	48	}
50		49
51	-#endif // SSYR2_GOLDEN_H

Mtest/syr2/ssyr2/ssyr2_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef SSYR2_PARAM_H	11	+#pragma once
12	-#define SSYR2_PARAM_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <string>	14	#include <string>
	@@ -43,4 +42,3 @@ struct Ssyr2Param : public BlasTestParamBase {
43	}	42	}
44	};	43	};
45		44
46	-#endif // SSYR2_PARAM_H

Mtest/tbmv/stbmv/arch35/stbmv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TBMV_NPU_H	11	+#pragma once
12	-#define TBMV_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <cstdlib>	14	#include <cstdlib>
	@@ -67,4 +66,3 @@ inline aclblasStatus_t aclblasStbmv_npu(
67	return ret;	66	return ret;
68	}	67	}
69		68
70	-#endif // TBMV_NPU_H

Mtest/tbmv/stbmv/stbmv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TBMV_CPU_H	11	+#pragma once
12	-#define TBMV_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -117,4 +116,3 @@ inline aclblasStatus_t aclblasStbmv_cpu(
117	return ACLBLAS_STATUS_SUCCESS;	116	return ACLBLAS_STATUS_SUCCESS;
118	}	117	}
119		118
120	-#endif // TBMV_CPU_H

Mtest/tbmv/stbmv/stbmv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TBMV_PARAM_H	11	+#pragma once
12	-#define TBMV_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -38,4 +37,3 @@ struct StbmvParam : public BlasTestParamBase {
38	}	37	}
39	};	38	};
40		39
41	-#endif // TBMV_PARAM_H

Mtest/tpmv/stpmv/arch35/stpmv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPMV_NPU_WRAPPER_H	11	+#pragma once
12	-#define STPMV_NPU_WRAPPER_H
13		12
14	#include <cstdlib>	13	#include <cstdlib>
15		14
	@@ -101,4 +100,3 @@ inline aclblasStatus_t aclblasStpmv_npu(
101	return ret;	100	return ret;
102	}	101	}
103		102
104	-#endif // STPMV_NPU_WRAPPER_H

Mtest/tpmv/stpmv/stpmv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPMV_GOLDEN_H	11	+#pragma once
12	-#define STPMV_GOLDEN_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <cstdlib>	14	#include <cstdlib>
	@@ -44,4 +43,3 @@ inline aclblasStatus_t aclblasStpmv_cpu(
44	return ACLBLAS_STATUS_SUCCESS;	43	return ACLBLAS_STATUS_SUCCESS;
45	}	44	}
46		45
47	-#endif // STPMV_GOLDEN_H

Mtest/tpmv/stpmv/stpmv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPMV_PARAM_H	11	+#pragma once
12	-#define STPMV_PARAM_H
13		12
14	#include "cann_ops_blas.h"	13	#include "cann_ops_blas.h"
15	#include "csv_loader.h"	14	#include "csv_loader.h"
	@@ -39,4 +38,3 @@ struct StpmvParam : public BlasTestParamBase {
39	}	38	}
40	};	39	};
41		40
42	-#endif // STPMV_PARAM_H

Mtest/tpsv/stpsv/arch35/stpsv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TPSV_NPU_H	11	+#pragma once
12	-#define TPSV_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -66,4 +65,3 @@ inline aclblasStatus_t aclblasStpsv_npu(
66	return ret;	65	return ret;
67	}	66	}
68		67
69	-#endif // TPSV_NPU_H

Mtest/tpsv/stpsv/stpsv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TPSV_CPU_H	11	+#pragma once
12	-#define TPSV_CPU_H
13		12
14	#include <cstddef>	13	#include <cstddef>
15	#include <cstdint>	14	#include <cstdint>
	@@ -111,4 +110,3 @@ inline aclblasStatus_t aclblasStpsv_cpu(
111	return ACLBLAS_STATUS_SUCCESS;	110	return ACLBLAS_STATUS_SUCCESS;
112	}	111	}
113		112
114	-#endif // TPSV_CPU_H

Mtest/tpsv/stpsv/stpsv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TPSV_PARAM_H	11	+#pragma once
12	-#define TPSV_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -34,4 +33,3 @@ struct TpsvParam : public BlasTestParamBase {
34	}	33	}
35	};	34	};
36		35
37	-#endif // TPSV_PARAM_H

Mtest/tpttr/stpttr/arch35/stpttr_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPTTR_NPU_H	11	+#pragma once
12	-#define STPTTR_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -65,4 +64,3 @@ inline aclblasStatus_t aclblasStpttr_npu(
65	return ret;	64	return ret;
66	}	65	}
67		66
68	-#endif // STPTTR_NPU_H

Mtest/tpttr/stpttr/stpttr_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPTTR_CPU_H	11	+#pragma once
12	-#define STPTTR_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -52,4 +51,3 @@ inline aclblasStatus_t aclblasStpttr_cpu(
52	return ACLBLAS_STATUS_SUCCESS;	51	return ACLBLAS_STATUS_SUCCESS;
53	}	52	}
54		53
55	-#endif // STPTTR_CPU_H

Mtest/tpttr/stpttr/stpttr_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STPTTR_PARAM_H	11	+#pragma once
12	-#define STPTTR_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -34,4 +33,3 @@ struct StpttrParam : public BlasTestParamBase {
34	}	33	}
35	};	34	};
36		35
37	-#endif // STPTTR_PARAM_H

Mtest/trmv/strmv/arch35/strmv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRMV_NPU_H	11	+#pragma once
12	-#define TRMV_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -66,4 +65,3 @@ inline aclblasStatus_t aclblasStrmv_npu(
66	return ret;	65	return ret;
67	}	66	}
68		67
69	-#endif // TRMV_NPU_H

Mtest/trmv/strmv/strmv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRMV_CPU_H	11	+#pragma once
12	-#define TRMV_CPU_H
13		12
14	#include <cmath>	13	#include <cmath>
15	#include <cstdint>	14	#include <cstdint>
	@@ -83,4 +82,3 @@ inline aclblasStatus_t aclblasStrmv_cpu(
83	return ACLBLAS_STATUS_SUCCESS;	82	return ACLBLAS_STATUS_SUCCESS;
84	}	83	}
85		84
86	-#endif // TRMV_CPU_H

Mtest/trmv/strmv/strmv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRMV_PARAM_H	11	+#pragma once
12	-#define TRMV_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -36,4 +35,3 @@ struct StrmvParam : public BlasTestParamBase {
36	}	35	}
37	};	36	};
38		37
39	-#endif // TRMV_PARAM_H

Mtest/trsv/strsv/arch22/strsv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRSV_NPU_H	11	+#pragma once
12	-#define TRSV_NPU_H
13		12
14	#include "acl/acl.h"	13	#include "acl/acl.h"
15	#include "cann_ops_blas.h"	14	#include "cann_ops_blas.h"
	@@ -31,4 +30,3 @@ inline aclblasStatus_t aclblasStrsv_npu(
31	return aclblasStrsv(handle, uplo, trans, diag, n, A, lda, x, incx);	30	return aclblasStrsv(handle, uplo, trans, diag, n, A, lda, x, incx);
32	}	31	}
33		32
34	-#endif // TRSV_NPU_H

Mtest/trsv/strsv/arch35/strsv_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRSV_NPU_H	11	+#pragma once
12	-#define TRSV_NPU_H
13		12
14	#include "acl/acl.h"	13	#include "acl/acl.h"
15	#include "cann_ops_blas.h"	14	#include "cann_ops_blas.h"
	@@ -31,4 +30,3 @@ inline aclblasStatus_t aclblasStrsv_npu(
31	return aclblasStrsv(handle, uplo, trans, diag, n, A, lda, x, incx);	30	return aclblasStrsv(handle, uplo, trans, diag, n, A, lda, x, incx);
32	}	31	}
33		32
34	-#endif // TRSV_NPU_H

Mtest/trsv/strsv/trsv_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRSV_CPU_H	11	+#pragma once
12	-#define TRSV_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cmath>	14	#include <cmath>
	@@ -94,4 +93,3 @@ inline aclblasStatus_t aclblasStrsv_cpu(
94	return ACLBLAS_STATUS_SUCCESS;	93	return ACLBLAS_STATUS_SUCCESS;
95	}	94	}
96		95
97	-#endif // TRSV_CPU_H

Mtest/trsv/strsv/trsv_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef TRSV_PARAM_H	11	+#pragma once
12	-#define TRSV_PARAM_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15	#include <string>	14	#include <string>
	@@ -41,4 +40,3 @@ struct TrsvParam : public BlasTestParamBase {
41	}	40	}
42	};	41	};
43		42
44	-#endif // TRSV_PARAM_H

Mtest/trttp/strttp/arch35/strttp_npu_wrapper.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STRTTP_NPU_H	11	+#pragma once
12	-#define STRTTP_NPU_H
13		12
14	#include <cstdint>	13	#include <cstdint>
15		14
	@@ -63,4 +62,3 @@ inline aclblasStatus_t aclblasStrttp_npu(
63	return ret;	62	return ret;
64	}	63	}
65		64
66	-#endif // STRTTP_NPU_H

Mtest/trttp/strttp/strttp_golden.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STRTTP_CPU_H	11	+#pragma once
12	-#define STRTTP_CPU_H
13		12
14	#include <algorithm>	13	#include <algorithm>
15	#include <cstdint>	14	#include <cstdint>
	@@ -45,4 +44,3 @@ inline aclblasStatus_t aclblasStrttp_cpu(
45	return ACLBLAS_STATUS_SUCCESS;	44	return ACLBLAS_STATUS_SUCCESS;
46	}	45	}
47		46
48	-#endif // STRTTP_CPU_H

Mtest/trttp/strttp/strttp_param.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef STRTTP_PARAM_H	11	+#pragma once
12	-#define STRTTP_PARAM_H
13		12
14	#include <string>	13	#include <string>
15		14
	@@ -34,4 +33,3 @@ struct StrttpParam : public BlasTestParamBase {
34	}	33	}
35	};	34	};
36		35
37	-#endif // STRTTP_PARAM_H

Mtest/utils/cblas_compat.h+1-3

	@@ -8,8 +8,7 @@
8	* See LICENSE in the root of the software repository for the full text of the License.	8	* See LICENSE in the root of the software repository for the full text of the License.
9	*/	9	*/
10		10
11	-#ifndef OPS_BLAS_TEST_UTILS_CBLAS_COMPAT_H	11	+#pragma once
12	-#define OPS_BLAS_TEST_UTILS_CBLAS_COMPAT_H
13		12
14	#include <cassert>	13	#include <cassert>
15		14
	@@ -71,4 +70,3 @@ inline CBLAS_DIAG ToCblasDiag(aclblasDiagType_t diag)
71	}	70	}
72	}	71	}
73		72
74	-#endif // OPS_BLAS_TEST_UTILS_CBLAS_COMPAT_H

Mtest/utils/error_check.h+1-3

	@@ -1,5 +1,4 @@
1	-#ifndef OPS_BLAS_TEST_UTILS_ERROR_CHECK_H	1	+#pragma once
2	-#define OPS_BLAS_TEST_UTILS_ERROR_CHECK_H
3		2
4	#include <cstdlib>	3	#include <cstdlib>
5	#include <cstdio>	4	#include <cstdio>
	@@ -38,4 +37,3 @@
38	} \	37	} \
39	}	38	}
40		39
41	-#endif // OPS_BLAS_TEST_UTILS_ERROR_CHECK_H

Mtest/utils/golden.h+1-3

	@@ -1,5 +1,4 @@
1	-#ifndef OPS_BLAS_TEST_UTILS_GOLDEN_H	1	+#pragma once
2	-#define OPS_BLAS_TEST_UTILS_GOLDEN_H
3		2
4	#include <algorithm>	3	#include <algorithm>
5	#include <cmath>	4	#include <cmath>
	@@ -92,4 +91,3 @@ inline void parseArguments(int argc, char* argv[], int& m, int& k, int& n)
92	}	91	}
93	}	92	}
94		93
95	-#endif // OPS_BLAS_TEST_UTILS_GOLDEN_H