diff options
| author | wuzhenqing <[email protected]> | 2026-05-04 11:15:50 +0000 |
|---|---|---|
| committer | wuzhenqing <[email protected]> | 2026-05-04 11:18:09 +0000 |
| commit | fa807b737b4b6b3004cccafc0dfdfdd66fcdce2d (patch) | |
| tree | 558dc6bc9e5749180fa97a569aab8bb6c2e54fec /tests/projects/ascendc/mixed_binary | |
| parent | 8bbbb2860fdd1910db466fe318ee8d465728ca32 (diff) | |
feat(ascendc): add Huawei Ascend C toolchain support
Add toolchain, language, rules, and compiler modules for the
bisheng compiler driver from CANN (Compute Architecture for
Neural Networks) toolkit. Supports .asc (Ascend C kernel) and
.aicpu (AI-CPU operator) source kinds with binary/static/shared
target types.
- Toolchain: SDK auto-detection via ASCEND_HOME_PATH env or --sdk
- Compiler: bisheng module inherits from clang, adds add_sourceflags
for explicit sourcekind overrides
- Rules: ascendc.env (default C++17), npuarchs (--npu-arch), build
- Language: check_main detection, add_ascnpuarchs API
- Tests: unit tests for check_main, project tests for libs/mixed
Fixes #7506
Diffstat (limited to 'tests/projects/ascendc/mixed_binary')
| -rw-r--r-- | tests/projects/ascendc/mixed_binary/src/helper.aicpu | 6 | ||||
| -rw-r--r-- | tests/projects/ascendc/mixed_binary/src/main.asc | 25 | ||||
| -rw-r--r-- | tests/projects/ascendc/mixed_binary/test.lua | 9 | ||||
| -rw-r--r-- | tests/projects/ascendc/mixed_binary/xmake.lua | 6 |
4 files changed, 46 insertions, 0 deletions
diff --git a/tests/projects/ascendc/mixed_binary/src/helper.aicpu b/tests/projects/ascendc/mixed_binary/src/helper.aicpu new file mode 100644 index 000000000..a11afdd90 --- /dev/null +++ b/tests/projects/ascendc/mixed_binary/src/helper.aicpu @@ -0,0 +1,6 @@ +#include "aicpu_api.h" + +__global__ __aicpu__ uint32_t hello_world(void *args) { + AscendC::printf("Hello World!!!\n"); + return 0; +} diff --git a/tests/projects/ascendc/mixed_binary/src/main.asc b/tests/projects/ascendc/mixed_binary/src/main.asc new file mode 100644 index 000000000..9bc0999bd --- /dev/null +++ b/tests/projects/ascendc/mixed_binary/src/main.asc @@ -0,0 +1,25 @@ +#include "acl/acl.h" + +struct KernelArgs { + int mode; +}; + +extern __global__ __aicpu__ uint32_t hello_world(void *args); + +int32_t main(int argc, char const *argv[]) { + aclInit(nullptr); + int32_t deviceId = 0; + aclrtSetDevice(deviceId); + aclrtStream stream = nullptr; + aclrtCreateStream(&stream); + + struct KernelArgs args = {0}; + constexpr uint32_t numBlocks = 1; + hello_world<<<numBlocks, nullptr, stream>>>(&args, sizeof(KernelArgs)); + aclrtSynchronizeStream(stream); + + aclrtDestroyStream(stream); + aclrtResetDevice(deviceId); + aclFinalize(); + return 0; +} diff --git a/tests/projects/ascendc/mixed_binary/test.lua b/tests/projects/ascendc/mixed_binary/test.lua new file mode 100644 index 000000000..0a36e62e4 --- /dev/null +++ b/tests/projects/ascendc/mixed_binary/test.lua @@ -0,0 +1,9 @@ +function main(t) + if not is_host("linux") then + return t:skip("wrong host platform") + end + if not os.getenv("ASCEND_HOME_PATH") then + return t:skip("ASCEND_HOME_PATH not set") + end + t:build() +end diff --git a/tests/projects/ascendc/mixed_binary/xmake.lua b/tests/projects/ascendc/mixed_binary/xmake.lua new file mode 100644 index 000000000..046da8404 --- /dev/null +++ b/tests/projects/ascendc/mixed_binary/xmake.lua @@ -0,0 +1,6 @@ +add_rules("mode.debug", "mode.release") + +target("ascendc_mixed") + set_kind("binary") + add_files("src/main.asc", "src/helper.aicpu") + add_ascnpuarchs("dav-2201") |
