From 21f14055b1f550ab12ac549ed089457ed307dcd1 Mon Sep 17 00:00:00 2001 From: JackLau1222 <2366536135@qq.com> Date: Wed, 24 Apr 2024 16:55:15 +0800 Subject: [PATCH 1/2] update the HelloWorldSample to support 310B4 and more core version --- operator/HelloWorldSample/CMakeLists.txt | 44 ++---- .../cmake/Modules/CMakeCCECompiler.cmake.in | 5 + .../cmake/Modules/CMakeCCEInformation.cmake | 41 ++++++ .../Modules/CMakeDetermineCCECompiler.cmake | 124 +++++++++++++++++ .../cmake/Modules/CMakeTestCCECompiler.cmake | 1 + .../HelloWorldSample/cmake/cpu/CMakeLists.txt | 37 +++++ .../HelloWorldSample/cmake/npu/CMakeLists.txt | 26 ++++ operator/HelloWorldSample/hello_world.cpp | 4 +- operator/HelloWorldSample/main.cpp | 23 +++- operator/HelloWorldSample/run.sh | 127 +++++++++++++++++- 10 files changed, 386 insertions(+), 46 deletions(-) create mode 100644 operator/HelloWorldSample/cmake/Modules/CMakeCCECompiler.cmake.in create mode 100644 operator/HelloWorldSample/cmake/Modules/CMakeCCEInformation.cmake create mode 100755 operator/HelloWorldSample/cmake/Modules/CMakeDetermineCCECompiler.cmake create mode 100644 operator/HelloWorldSample/cmake/Modules/CMakeTestCCECompiler.cmake create mode 100644 operator/HelloWorldSample/cmake/cpu/CMakeLists.txt create mode 100644 operator/HelloWorldSample/cmake/npu/CMakeLists.txt mode change 100755 => 100644 operator/HelloWorldSample/run.sh diff --git a/operator/HelloWorldSample/CMakeLists.txt b/operator/HelloWorldSample/CMakeLists.txt index 6ee57bb82..fd87c7620 100644 --- a/operator/HelloWorldSample/CMakeLists.txt +++ b/operator/HelloWorldSample/CMakeLists.txt @@ -1,38 +1,10 @@ -# Copyright (c) Huawei Technologies Co., Ltd. 2020. All rights reserved. +cmake_minimum_required(VERSION 3.16) +set(CMAKE_SYSTEM_NAME Linux) +set(CMAKE_CXX_STANDARD 17) -# CMake lowest version requirement -cmake_minimum_required(VERSION 3.16.0) +set(CCE_CMAKE_PATH ${CMAKE_CURRENT_SOURCE_DIR}/cmake/Modules) +list(APPEND CMAKE_MODULE_PATH ${CCE_CMAKE_PATH}) +project(kernel_samples LANGUAGES CCE CXX) -# project information -project(Ascend_C) -set(SOC_VERSION "Ascend910B1" CACHE STRING "system on chip type") -set(ASCEND_CANN_PACKAGE_PATH "~/Ascend/ascend-toolkit/latest" CACHE PATH "ASCEND CANN package installation directory") -set(RUN_MODE "npu" CACHE STRING "run mode: npu/sim/cpu") -set(CMAKE_BUILD_TYPE "Debug" CACHE STRING "Build type Release/Debug (default Debug)" FORCE) -set(CMAKE_INSTALL_PREFIX "${CMAKE_CURRENT_LIST_DIR}/out" CACHE STRING "path for install()" FORCE) - -if(EXISTS ${ASCEND_CANN_PACKAGE_PATH}/tools/tikcpp/ascendc_kernel_cmake) - set(ASCENDC_CMAKE_DIR ${ASCEND_CANN_PACKAGE_PATH}/tools/tikcpp/ascendc_kernel_cmake) -elseif(EXISTS ${ASCEND_CANN_PACKAGE_PATH}/compiler/tikcpp/ascendc_kernel_cmake) - set(ASCENDC_CMAKE_DIR ${ASCEND_CANN_PACKAGE_PATH}/compiler/tikcpp/ascendc_kernel_cmake) -elseif(EXISTS ${ASCEND_CANN_PACKAGE_PATH}/ascendc_devkit/tikcpp/samples/cmake) - set(ASCENDC_CMAKE_DIR ${ASCEND_CANN_PACKAGE_PATH}/ascendc_devkit/tikcpp/samples/cmake) -else() - message(FATAL_ERROR "ascendc_kernel_cmake does not exist, please check whether the cann package is installed.") -endif() - -include(${ASCENDC_CMAKE_DIR}/ascendc.cmake) - -ascendc_library(kernels STATIC - hello_world.cpp -) - -ascendc_compile_definitions(kernels PRIVATE - -DASCENDC_DUMP -) - -add_executable(main main.cpp) - -target_link_libraries(main PRIVATE - kernels -) \ No newline at end of file +add_subdirectory(cmake/cpu) +add_subdirectory(cmake/npu) \ No newline at end of file diff --git a/operator/HelloWorldSample/cmake/Modules/CMakeCCECompiler.cmake.in b/operator/HelloWorldSample/cmake/Modules/CMakeCCECompiler.cmake.in new file mode 100644 index 000000000..a9b5688ff --- /dev/null +++ b/operator/HelloWorldSample/cmake/Modules/CMakeCCECompiler.cmake.in @@ -0,0 +1,5 @@ +set(CMAKE_CCE_COMPILER "@CMAKE_CCE_COMPILER@") +set(CMAKE_CCE_COMPILER_LOADED 1) +set(CMAKE_CCE_SOURCE_FILE_EXTENSIONS @CMAKE_CCE_SOURCE_FILE_EXTENSIONS@) +set(CMAKE_CCE_OUTPUT_EXTENSION @CMAKE_CCE_OUTPUT_EXTENSION@) +set(CMAKE_CCE_COMPILER_ENV_VAR "@CMAKE_CCE_COMPILER_ENV_VAR@") diff --git a/operator/HelloWorldSample/cmake/Modules/CMakeCCEInformation.cmake b/operator/HelloWorldSample/cmake/Modules/CMakeCCEInformation.cmake new file mode 100644 index 000000000..7d2fc2ddd --- /dev/null +++ b/operator/HelloWorldSample/cmake/Modules/CMakeCCEInformation.cmake @@ -0,0 +1,41 @@ +include(CMakeCommonLanguageInclude) + +set(CMAKE_INCLUDE_FLAG_CCE "-I") + +if(UNIX) + set(CMAKE_CCE_OUTPUT_EXTENSION .o) +else() + set(CMAKE_CCE_OUTPUT_EXTENSION .obj) +endif() + +set(_INCLUDED_FILE 0) +set(CMAKE_SHARED_LIBRARY_CCE_FLAGS -fPIC) +set(CMAKE_SHARED_LIBRARY_CREATE_CCE_FLAGS -shared) +set(CMAKE_STATIC_LIBRARY_CREATE_CCE_FLAGS "--cce-build-static-lib") +set(CMAKE_LIBRARY_CREATE_CCE_FLAGS "--cce-fatobj-link") + +if(NOT CMAKE_CCE_COMPILE_OBJECT) + set(CMAKE_CCE_COMPILE_OBJECT + " -xcce ${__IMPLICIT_INCLUDES} ${_CMAKE_CCE_BUILTIN_INCLUDE_PATH} ${_CMAKE_COMPILE_AS_CCE_FLAG} ${_CMAKE_CCE_COMPILE_OPTIONS} ${_CMAKE_CCE_COMMON_COMPILE_OPTIONS} -pthread -o -c ") +endif() + +if(NOT CMAKE_CCE_CREATE_SHARED_LIBRARY) + set(CMAKE_CCE_CREATE_SHARED_LIBRARY + " ${CMAKE_LIBRARY_CREATE_CCE_FLAGS} -o ") +endif() + +if(NOT CMAKE_CCE_CREATE_STATIC_LIBRARY) + set(CMAKE_CCE_CREATE_STATIC_LIBRARY + " ${CMAKE_LIBRARY_CREATE_CCE_FLAGS} -o ") +endif() + +if(NOT CMAKE_CCE_CREATE_SHARED_MODULE) + set(CMAKE_CCE_CREATE_SHARED_MODULE ${CMAKE_CCE_CREATE_SHARED_LIBRARY}) +endif() + +if(NOT CMAKE_CCE_LINK_EXECUTABLE) + set(CMAKE_CCE_LINK_EXECUTABLE + " ${CMAKE_LIBRARY_CREATE_CCE_FLAGS} -o ${__IMPLICIT_LINKS}") +endif() + +set(CMAKE_CCE_INFORMATION_LOADED 1) diff --git a/operator/HelloWorldSample/cmake/Modules/CMakeDetermineCCECompiler.cmake b/operator/HelloWorldSample/cmake/Modules/CMakeDetermineCCECompiler.cmake new file mode 100755 index 000000000..f3c0a4732 --- /dev/null +++ b/operator/HelloWorldSample/cmake/Modules/CMakeDetermineCCECompiler.cmake @@ -0,0 +1,124 @@ +find_program(CMAKE_CCE_COMPILER NAMES "ccec" PATHS "$ENV{PATH}" DOC "CCE Compiler") + +mark_as_advanced(CMAKE_CCE_COMPILER) + +message(STATUS "CMAKE_CCE_COMPILER: " ${CMAKE_CCE_COMPILER}) +set(CMAKE_CCE_SOURCE_FILE_EXTENSIONS cce;cpp) +set(CMAKE_CCE_COMPILER_ENV_VAR "CCE") +message(STATUS "CMAKE_CURRENT_LIST_DIR: " ${CMAKE_CURRENT_LIST_DIR}) + +# configure all variables set in this file +configure_file(${CMAKE_CURRENT_LIST_DIR}/CMakeCCECompiler.cmake.in + ${CMAKE_PLATFORM_INFO_DIR}/CMakeCCECompiler.cmake + @ONLY +) + +message(STATUS "ASCEND_PRODUCT_TYPE:\n" " ${ASCEND_PRODUCT_TYPE}") +message(STATUS "ASCEND_CORE_TYPE:\n" " ${ASCEND_CORE_TYPE}") +message(STATUS "ASCEND_INSTALL_PATH:\n" " ${ASCEND_INSTALL_PATH}") + +if(DEFINED ASCEND_INSTALL_PATH) + set(_CMAKE_ASCEND_INSTALL_PATH ${ASCEND_INSTALL_PATH}) +else() + message(FATAL_ERROR + "no, installation path found, should passing -DASCEND_INSTALL_PATH= in cmake" + ) + set(_CMAKE_ASCEND_INSTALL_PATH) +endif() + + +if(DEFINED ASCEND_PRODUCT_TYPE) + set(_CMAKE_CCE_COMMON_COMPILE_OPTIONS "--cce-auto-sync -mllvm -api-deps-filter") + if(ASCEND_PRODUCT_TYPE STREQUAL "") + message(FATAL_ERROR "ASCEND_PRODUCT_TYPE must be non-empty if set.") + elseif(ASCEND_PRODUCT_TYPE AND NOT ASCEND_PRODUCT_TYPE MATCHES "^Ascend[0-9][0-9][0-9][a-zA-Z]?[1-9]?$") + message(FATAL_ERROR + "ASCEND_PRODUCT_TYPE: ${ASCEND_PRODUCT_TYPE}\n" + "is not one of the following: Ascend910A, Ascend310P1, Ascend910B1, Ascend310B1" + ) + elseif(ASCEND_PRODUCT_TYPE STREQUAL "Ascend910A") + if (ASCEND_CORE_TYPE STREQUAL "AiCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-c100") + else() + message(FATAL_ERROR, "only AiCore inside") + endif() + set(_CMAKE_CCE_COMPILE_OPTIONS) + elseif(ASCEND_PRODUCT_TYPE STREQUAL "Ascend310P1") + if (ASCEND_CORE_TYPE STREQUAL "AiCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-m200") + elseif(ASCEND_CORE_TYPE STREQUAL "VectorCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-m200-vec") + endif() + set(_CMAKE_CCE_COMPILE_OPTIONS + "-mllvm -cce-aicore-function-stack-size=16000 -mllvm -cce-aicore-fp-ceiling=2 -mllvm -cce-aicore-record-overflow=false") + elseif(ASCEND_PRODUCT_TYPE STREQUAL "Ascend310B1") + if (ASCEND_CORE_TYPE STREQUAL "AiCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-m300") + elseif(ASCEND_CORE_TYPE STREQUAL "VectorCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-m300-vec") + endif() + set(_CMAKE_CCE_COMPILE_OPTIONS + "-mllvm -cce-aicore-function-stack-size=16000 -mllvm -cce-aicore-fp-ceiling=2 -mllvm -cce-aicore-record-overflow=false") + elseif(ASCEND_PRODUCT_TYPE STREQUAL "Ascend910B1") + if (ASCEND_CORE_TYPE STREQUAL "AiCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-c220-cube") + elseif(ASCEND_CORE_TYPE STREQUAL "VectorCore") + set(_CMAKE_COMPILE_AS_CCE_FLAG "--cce-aicore-arch=dav-c220-vec") + endif() + set(_CMAKE_CCE_COMPILE_OPTIONS + "-mllvm -cce-aicore-function-stack-size=16000 -mllvm -cce-aicore-record-overflow=false -mllvm -cce-aicore-addr-transform") + endif() +endif() + +set(_CMAKE_CCE_HOST_IMPLICIT_LINK_DIRECTORIES + ${_CMAKE_ASCEND_INSTALL_PATH}/runtime/lib64 + ${_CMAKE_ASCEND_INSTALL_PATH}/tools/simulator/${ASCEND_PRODUCT_TYPE}/lib + ${_CMAKE_ASCEND_INSTALL_PATH}/tools/tikicpulib/lib/${ASCEND_PRODUCT_TYPE} +) + +# link library +set(_CMAKE_CCE_HOST_IMPLICIT_LINK_LIBRARIES stdc++) +if(ASCEND_RUN_MODE STREQUAL "npu") + list(APPEND _CMAKE_CCE_HOST_IMPLICIT_LINK_LIBRARIES runtime) +elseif(ASCEND_RUN_MODE STREQUAL "sim") + list(APPEND _CMAKE_CCE_HOST_IMPLICIT_LINK_DIRECTORIES ) + if(ASCEND_PRODUCT_TYPE STREQUAL "Ascend910A") + list(APPEND _CMAKE_CCE_HOST_IMPLICIT_LINK_LIBRARIES pem_davinci) + endif() + list(APPEND _CMAKE_CCE_HOST_IMPLICIT_LINK_LIBRARIES runtime_camodel) +elseif(ASCEND_RUN_MODE STREQUAL "cpu") + message(STATUS "RUN_MODE is cpu") +else() + message(FATAL_ERROR + "ASCEND_RUN_MODE: ${ASCEND_RUN_MODE}\n" + "ASCEND_RUN_MODE must be one of the following: cpu, npu or sim" + ) +endif() +list(APPEND _CMAKE_CCE_HOST_IMPLICIT_LINK_LIBRARIES ascendcl) + +set(__IMPLICIT_LINKS) +foreach(dir ${_CMAKE_CCE_HOST_IMPLICIT_LINK_DIRECTORIES}) + string(APPEND __IMPLICIT_LINKS " -L\"${dir}\"") +endforeach() +foreach(lib ${_CMAKE_CCE_HOST_IMPLICIT_LINK_LIBRARIES}) + if(${lib} MATCHES "/") + string(APPEND __IMPLICIT_LINKS " \"${lib}\"") + else() + string(APPEND __IMPLICIT_LINKS " -l${lib}") + endif() +endforeach() + +set(_CMAKE_CCE_HOST_IMPLICIT_INCLUDE_DIRECTORIES + ${_CMAKE_ASCEND_INSTALL_PATH}/acllib/include + ${_CMAKE_ASCEND_INSTALL_PATH}/compiler/tikcpp/tikcfw + ${_CMAKE_ASCEND_INSTALL_PATH}/compiler/tikcpp/tikcfw/impl + ${_CMAKE_ASCEND_INSTALL_PATH}/compiler/tikcpp/tikcfw/interface + ${_CMAKE_ASCEND_INSTALL_PATH}/tools/tikcpp/tikcfw + ${_CMAKE_ASCEND_INSTALL_PATH}/tools/tikcpp/tikcfw/impl + ${_CMAKE_ASCEND_INSTALL_PATH}/tools/tikcpp/tikcfw/interface + ${_CMAKE_ASCEND_INSTALL_PATH}/tools/tikicpulib/lib/include +) +set(__IMPLICIT_INCLUDES) +foreach(inc ${_CMAKE_CCE_HOST_IMPLICIT_INCLUDE_DIRECTORIES}) + string(APPEND __IMPLICIT_INCLUDES " -I\"${inc}\"") +endforeach() \ No newline at end of file diff --git a/operator/HelloWorldSample/cmake/Modules/CMakeTestCCECompiler.cmake b/operator/HelloWorldSample/cmake/Modules/CMakeTestCCECompiler.cmake new file mode 100644 index 000000000..f00f227c1 --- /dev/null +++ b/operator/HelloWorldSample/cmake/Modules/CMakeTestCCECompiler.cmake @@ -0,0 +1 @@ +set(CMAKE_CCE_COMPILER_WORKS 1 CACHE INTERNAL "") diff --git a/operator/HelloWorldSample/cmake/cpu/CMakeLists.txt b/operator/HelloWorldSample/cmake/cpu/CMakeLists.txt new file mode 100644 index 000000000..661230b17 --- /dev/null +++ b/operator/HelloWorldSample/cmake/cpu/CMakeLists.txt @@ -0,0 +1,37 @@ +# cpu +if (NOT DEFINED ENV{CMAKE_PREFIX_PATH}) + set(CMAKE_PREFIX_PATH ${ASCEND_INSTALL_PATH}/tools/tikicpulib/lib/cmake) +endif() + +find_package(tikicpulib REQUIRED) + +file(GLOB SRC_FILES + ${CMAKE_SOURCE_DIR}/*.cpp +) + +add_executable(${smoke_testcase}_cpu + ${SRC_FILES} +) + +target_include_directories(${smoke_testcase}_cpu PRIVATE + ${ASCEND_INSTALL_PATH}/acllib/include + ${CMAKE_SOURCE_DIR} +) + +target_link_libraries(${smoke_testcase}_cpu PRIVATE + tikicpulib::${ASCEND_PRODUCT_TYPE} + ascendcl +) + +target_compile_options(${smoke_testcase}_cpu PRIVATE + -g +) + +target_compile_definitions(${smoke_testcase}_cpu PRIVATE + _GLIBCXX_USE_CXX11_ABI=0 +) + +set_target_properties(${smoke_testcase}_cpu PROPERTIES + OUTPUT_NAME ${smoke_testcase}_${ASCEND_RUN_MODE} + RUNTIME_OUTPUT_DIRECTORY ${CMAKE_SOURCE_DIR} +) diff --git a/operator/HelloWorldSample/cmake/npu/CMakeLists.txt b/operator/HelloWorldSample/cmake/npu/CMakeLists.txt new file mode 100644 index 000000000..fb1786c6b --- /dev/null +++ b/operator/HelloWorldSample/cmake/npu/CMakeLists.txt @@ -0,0 +1,26 @@ +# npu +file(GLOB SRC_FILES + ${CMAKE_SOURCE_DIR}/*.cpp +) +set_source_files_properties(${SRC_FILES} PROPERTIES LANGUAGE CCE) + +add_executable(${smoke_testcase}_npu + ${SRC_FILES} +) + +target_compile_options(${smoke_testcase}_npu PRIVATE + -O2 + -std=c++17 +) + +target_link_directories(${smoke_testcase}_npu PRIVATE + ${ASCEND_INSTALL_PATH}/lib64 +) + +set_target_properties(${smoke_testcase}_npu PROPERTIES + OUTPUT_NAME ${smoke_testcase}_${ASCEND_RUN_MODE} + RUNTIME_OUTPUT_DIRECTORY ${CMAKE_SOURCE_DIR} +) + +add_custom_target(${smoke_testcase}_sim ALL) +add_dependencies(${smoke_testcase}_sim ${smoke_testcase}_npu) diff --git a/operator/HelloWorldSample/hello_world.cpp b/operator/HelloWorldSample/hello_world.cpp index 66457bc10..41506b4a8 100644 --- a/operator/HelloWorldSample/hello_world.cpp +++ b/operator/HelloWorldSample/hello_world.cpp @@ -7,6 +7,8 @@ extern "C" __global__ __aicore__ void hello_world() { PRINTF("Hello World!!!\n"); } +#ifndef __CCE_KT_TEST__ void hello_world_do(uint32_t blockDim, void* stream) { hello_world<<>>(); -} \ No newline at end of file +} +#endif diff --git a/operator/HelloWorldSample/main.cpp b/operator/HelloWorldSample/main.cpp index 2fb27ae94..53a40f1c9 100644 --- a/operator/HelloWorldSample/main.cpp +++ b/operator/HelloWorldSample/main.cpp @@ -3,10 +3,22 @@ * This file constains code of cpu debug and npu code.We read data from bin file * and write result to file. */ +#include +#ifndef __CCE_KT_TEST__ #include "acl/acl.h" -extern void hello_world_do(uint32_t coreDim, void* stream); +void hello_world_do(uint32_t blockDim, void* stream); +#else +extern "C" __global__ __aicore__ void hello_world(); +#endif -int32_t main(int argc, char const *argv[]) { +int32_t main(int32_t argc, char* argv[]) +{ + uint32_t blockDim = 8; + +#ifdef __CCE_KT_TEST__ + AscendC::SetKernelMode(KernelMode::AIV_MODE); + ICPU_RUN_KF(hello_world, blockDim); // use this macro for cpu debug +#else aclInit(nullptr); aclrtContext context; int32_t deviceId = 0; @@ -14,14 +26,17 @@ int32_t main(int argc, char const *argv[]) { aclrtCreateContext(&context,deviceId); aclrtStream stream = nullptr; aclrtCreateStream(&stream); + + std::cout << "Use the npu: Hello World!\n"; + hello_world_do(8, stream); - constexpr uint32_t blockDim = 8; - hello_world_do(blockDim, stream); aclrtSynchronizeStream(stream); aclrtDestroyStream(stream); aclrtDestroyContext(context); aclrtResetDevice(deviceId); aclFinalize(); + +#endif return 0; } diff --git a/operator/HelloWorldSample/run.sh b/operator/HelloWorldSample/run.sh old mode 100755 new mode 100644 index 9d34a52eb..07097f4d9 --- a/operator/HelloWorldSample/run.sh +++ b/operator/HelloWorldSample/run.sh @@ -1,10 +1,127 @@ #!/bin/bash +export PRINT_TIK_MEM_ACCESS=FALSE + +CURRENT_DIR=$( + cd $(dirname ${BASH_SOURCE:-$0}) + pwd +); cd $CURRENT_DIR + +declare -A VersionMap +VersionMap["Ascend910A"]="Ascend910A" +VersionMap["Ascend910B"]="Ascend910A" +VersionMap["Ascend910ProA"]="Ascend910A" +VersionMap["Ascend910ProB"]="Ascend910A" +VersionMap["Ascend910PremiumA"]="Ascend910A" +VersionMap["Ascend310B1"]="Ascend310B1" +VersionMap["Ascend310B2"]="Ascend310B1" +VersionMap["Ascend310B3"]="Ascend310B1" +VersionMap["Ascend310B4"]="Ascend310B1" +VersionMap["Ascend310P1"]="Ascend310P1" +VersionMap["Ascend310P3"]="Ascend310P1" +VersionMap["Ascend910B1"]="Ascend910B1" +VersionMap["Ascend910B2"]="Ascend910B1" +VersionMap["Ascend910B3"]="Ascend910B1" +VersionMap["Ascend910B4"]="Ascend910B1" +# legacy +VersionMap["ascend910"]="Ascend910A" +VersionMap["ascend310p"]="Ascend310P1" +VersionMap["ascend310B1"]="Ascend310B1" +VersionMap["ascend910B1"]="Ascend910B1" + +FILE_NAME="hello_world" + +SHORT=r:,v:,i:, +LONG=run-mode:,soc-version:,install-path:, +OPTS=$(getopt -a --options $SHORT --longoptions $LONG -- "$@") +eval set -- "$OPTS" + +while : +do + case "$1" in + (-r | --run-mode ) + RUN_MODE="$2" + shift 2;; + (-v | --soc-version ) + SOC_VERSION="$2" + shift 2;; + (-i | --install-path ) + ASCEND_INSTALL_PATH="$2" + shift 2;; + (--) + shift; + break;; + (*) + echo "[ERROR] Unexpected option: $1"; + break;; + esac +done + + +if [ -n "$ASCEND_INSTALL_PATH" ]; then + _ASCEND_INSTALL_PATH=$ASCEND_INSTALL_PATH +elif [ -n "$ASCEND_HOME_PATH" ]; then + _ASCEND_INSTALL_PATH=$ASCEND_HOME_PATH +else + if [ -d "$HOME/Ascend/ascend-toolkit/latest" ]; then + _ASCEND_INSTALL_PATH=$HOME/Ascend/ascend-toolkit/latest + else + _ASCEND_INSTALL_PATH=/usr/local/Ascend/ascend-toolkit/latest + fi +fi +# in case of running op in simulator, use stub so instead +if [ "${RUN_MODE}" = "sim" ]; then + export LD_LIBRARY_PATH=$_ASCEND_INSTALL_PATH/runtime/lib64/stub:$LD_LIBRARY_PATH + if [ ! $CAMODEL_LOG_PATH ]; then + export CAMODEL_LOG_PATH=./sim_log + fi + rm -rf $CAMODEL_LOG_PATH + mkdir -p $CAMODEL_LOG_PATH +fi +source $_ASCEND_INSTALL_PATH/bin/setenv.bash + +if [[ " ${!VersionMap[*]} " != *" $SOC_VERSION "* ]]; then + echo "ERROR: SOC_VERSION should be in [${!VersionMap[*]}]" + exit -1 +fi +_SOC_VERSION=${VersionMap[$SOC_VERSION]} + +if [ $_SOC_VERSION"x" = "Ascend910Ax" ] || [ $_SOC_VERSION"x" = "Ascend310P1x" ] || [ $_SOC_VERSION"x" = "Ascend310B1x" ]; then + CORE_TYPE="AiCore" +elif [ $_SOC_VERSION"x" = "Ascend910B1x" ]; then + CORE_TYPE="VectorCore" +fi + +RUN_MODE_LIST="cpu sim npu" +if [[ " $RUN_MODE_LIST " != *" $RUN_MODE "* ]]; then + echo "ERROR: RUN_MODE error, This sample only support specify cpu, sim or npu!" + exit -1 +fi set -e -rm -rf build +rm -rf build *_cpu *_sim *_npu cceprint npuchk *log *.vcd + mkdir build -cmake -B build -cmake --build build -j -cmake --install build -./build/main +cmake -B build \ + -Dsmoke_testcase=${FILE_NAME} \ + -DASCEND_PRODUCT_TYPE=${_SOC_VERSION} \ + -DASCEND_CORE_TYPE=${CORE_TYPE} \ + -DASCEND_RUN_MODE=${RUN_MODE} \ + -DASCEND_INSTALL_PATH=${_ASCEND_INSTALL_PATH} +cmake --build build --target ${FILE_NAME}_${RUN_MODE} +if [ $? -ne 0 ]; then + echo "ERROR: compile op on failed!" + exit -1 +fi +echo "INFO: compile op on ${RUN_MODE} succeed!" + +#rm -rf input/*.bin output/*.bin +#python3 scripts/gen_data.py +(export LD_LIBRARY_PATH=${_ASCEND_INSTALL_PATH}/tools/simulator/${_SOC_VERSION}/lib:$LD_LIBRARY_PATH && ./${FILE_NAME}_${RUN_MODE}) +if [ $? -ne 0 ]; then + echo "ERROR: execute op on ${RUN_MODE} failed!" + exit -1 +fi +echo "INFO: execute op on ${RUN_MODE} succeed!" +#python3 scripts/verify_result.py output/output_z.bin output/golden.bin +rm -rf *log *.vcd -- Gitee From 14fb16258e7670574e79497c2ca816b01f76dfa0 Mon Sep 17 00:00:00 2001 From: JackLau1222 <2366536135@qq.com> Date: Wed, 24 Apr 2024 17:07:47 +0800 Subject: [PATCH 2/2] update README.md --- operator/README.md | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/operator/README.md b/operator/README.md index 3067dbb84..9fd9cc6ec 100644 --- a/operator/README.md +++ b/operator/README.md @@ -3,7 +3,7 @@ | ------------------------------------------------------------ | ---------------------------------------------------- | -- | | [AddcdivCustomSample](./AddcdivCustomSample) | 基于Ascend C的Addcdiv自定义Vector算子及调用样例 | Atlas训练系列产品
Atlas推理系列产品(Ascend 310P 处理器)
Atlas A2训练系列产品 | | [AddCustomSample](./AddCustomSample) | 基于Ascend C的Add自定义Vector算子及调用样例 | Atlas训练系列产品
Atlas 200/500 A2 推理产品
Atlas推理系列产品(Ascend 310P 处理器)
Atlas A2训练系列产品 | -| [HelloWorldSample](./HelloWorldSample) | 基于Ascend C的自定义算子调用结构演示样例 |Atlas训练系列产品
Atlas推理系列产品(Ascend 310P 处理器)
Atlas A2训练系列产品 | +| [HelloWorldSample](./HelloWorldSample) | 基于Ascend C的Add自定义Vector算子及调用样例 | Atlas训练系列产品
Atlas 200/500 A2 推理产品
Atlas推理系列产品(Ascend 310P 处理器)
Atlas A2训练系列产品 | | [LeakyReluCustomSample](./LeakyReluCustomSample) | 基于AscendC的LeakyReLU自定义Vector算子及调用样例 |Atlas推理系列产品(Ascend 310P 处理器)
Atlas 200/500 A2 推理产品
Atlas A2训练系列产品| | [LayerNormCustomSample](./LayerNormCustomSample) | 基于AscendC的LayerNorm自定义算子及调用样例 | Atlas A2训练系列产品 | | [MatMulCustomSample](./MatMulCustomSample) | 基于AscendC的Matmul自定义Cube算子及调用样例 | Atlas推理系列产品(Ascend 310P 处理器)
Atlas A2训练系列产品 | -- Gitee