Skip to main content

4. MUSA开发环境搭建及示例演示

请前往 MUSA SDK 下载 MUSA 最新版本。依赖环境如下：
Ubuntu 20.04.x LTS x86_64, Kernel version 5.4.0-42-generic

4.1. MUSA开发环境搭建​

4.1.1. 下载安装包​

目前安装包包含以下内容。

Category

Name

Package

musa_xorg_ubuntu2004_x86_20221125.deb

musa_toolkits_install_rc1.2.0.tar.gz

mudnn.tar

mttensorflow_whl.tar

Release Note

CUDA_Compatible_Release_Note_rc1.2.0.md

Installation Guide

Installation guide for Cuda compatible rc1.2.0.md

API DOCs

MUSA_API_rc1.2.0.pdf

MUSA_Compiler_Collection_rc1.2.0.pdf

MUSA_Runtime_API_rc1.2.0.pdf

Math-X_muBLAS_API_rc0.3.0.pdf

Math-X_muFFT_API_rc0.2.0.pdf

EULA

摩尔线程软件用户使用许可.docx

4.1.2. 安装教程​

4.1.2.1. 通用安装​

  1. 下载软件包
    > 敬请期待后续版本更新

  2. 安装KMD&UMD
    1）确保系统中已安装lightdm，安装命令如下：
    apt install lightdm （安装过程中提示configure lightdm，请务必选择lightdm，不要选择gdm3）
    2）cd ~/pkg
    3）dpkg -i musa_xorg_ubuntu2004_x86_20221125.deb
    4）modprobe mtgpu

  3. 安装MUSA Toolkits
    1）cd ~/pkg
    2）tar -zxvf musa_toolkits_install_rc1.2.0.tar.gz
    3）rm -rf /usr/local/musa*
    4）cd musa_toolkits_install
    5）./install.sh

NOTE: 如果需要支持AI的应用，则根据需求安装4和5

  1. 安装muDNN
    1）tar -xvf mudnn.tar
    2）cd mudnn && ./install_mudnn.sh -i

  2. 安装mtTensorFlow
    1）tar -xvf mttensorflow_whl.tar && cd mttensorflow_whl
    2）pip install tensorflow-2.8.2-cp38-cp38-linux_x86_64.whl
    (如果之前已安装TensorFlow，请先执行pip uninstall tensorflow)
    3）pip install mt_tf_plugin-0.0.1-cp38-cp38-linux_x86_64.whl
    (如果之前已安装TensorFlow，请先执行pip uninstall mt-tf-plugin)

  3. 添加 MUSA Toolkit的环境变量

    $ vim ~/.bashrc
    # 在其中加入以下内容
    export MUSA_INSTALL_PATH=/usr/local/musa
    export PATH=$MUSA_INSTALL_PATH/bin:$PATH
    export LD_LIBRARY_PATH=$MUSA_INSTALL_PATH/lib:$LD_LIBRARY_PATH
    # 完成以上内容
    $ source ~/.bashrc

4.2. MUSA示例演示​

4.2.1. 案例代码​

下面以一个简单的musa example：vector_add.mu参考,实现了浮点的vector相加。具体的代码如下：

#include <stdio.h>
#include <musa_runtime.h>

// musa kernel vector Add, C = A + B
__global__ void vectorAdd(const float *A, const float *B, float *C, int numElements) {
    int i = blockDim.x * blockIdx.x + threadIdx.x;
    if (i < numElements) {
        C[i] = A[i] + B[i];
    }
}

int main(void) {
    musaError_t err = musaSuccess;
    int numElements = 50000;
    size_t size = numElements * sizeof(float);
    float *h_A = (float *)malloc(size);
    float *h_B = (float *)malloc(size);
    float *h_C = (float *)malloc(size);
    if (h_A == NULL || h_B == NULL || h_C == NULL) {
        fprintf(stderr, "Failed to allocate host vectors!\n");
        exit(EXIT_FAILURE);
    }
    for (int i = 0; i < numElements; ++i) {
        h_A[i] = rand()/(float)RAND_MAX;
        h_B[i] = rand()/(float)RAND_MAX;
    }

    // Allocate the device input vector A
    float *d_A = NULL;
    err = musaMalloc((void **)&d_A, size);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to allocate device vector A (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    // Allocate the device input vector B
    float *d_B = NULL;
    err = musaMalloc((void **)&d_B, size);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to allocate device vector B (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    // Allocate the device output vector C
    float *d_C = NULL;
    err = musaMalloc((void **)&d_C, size);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to allocate device vector C (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }

    // Copy the host input vectors A and B in host memory to the device input vectors in
    // device memory
    printf("Copy input data from the host memory to the MUSA device\n");
    err = musaMemcpy(d_A, h_A, size, musaMemcpyHostToDevice);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to copy vector A from host to device (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    err = musaMemcpy(d_B, h_B, size, musaMemcpyHostToDevice);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to copy vector B from host to device (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    // Launch the Vector Add MUSA Kernel
    int threadsPerBlock = 256;
    int blocksPerGrid =(numElements + threadsPerBlock - 1) / threadsPerBlock;
    printf("MUSA kernel launch with %d blocks of %d threads\n", blocksPerGrid, threadsPerBlock);
    vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C, numElements);
    err = musaGetLastError();
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to launch vectorAdd kernel (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    // Copy the device result vector in device memory to the host result vector
    // in host memory.
    printf("Copy output data from the MUSA device to the host memory\n");
    err = musaMemcpy(h_C, d_C, size, musaMemcpyDeviceToHost);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to copy vector C from device to host (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    // Verify that the result vector is correct
    for (int i = 0; i < numElements; ++i) {
        if (fabs(h_A[i] + h_B[i] - h_C[i]) > 1e-5) {
            fprintf(stderr, "Result verification failed at element %d!\n", i);
            exit(EXIT_FAILURE);
        }
    }
    printf("Test PASSED\n");
    // Free device global memory
    err = musaFree(d_A);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to free device vector A (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    err = musaFree(d_B);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to free device vector B (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    err = musaFree(d_C);
    if (err != musaSuccess) {
        fprintf(stderr, "Failed to free device vector C (error code %s)!\n", musaGetErrorString(err));
        exit(EXIT_FAILURE);
    }
    // Free host memory
    free(h_A);
    free(h_B);
    free(h_C);
    printf("Done\n");
    return 0;
}

4.2.2. 编译并运行​

$ mcc vectorAdd.mu -o vectorAdd -lmusart
$ ./vectorAdd
Test PASSED
Done

4.3. 使用CMake构建MUSA项目​

cmake是一个构建软件工程的工具，可以在不同的平台上使用统一的脚本生成相应平台对应的Makefile或者build.ninja文件。自cmake版本2.8之后，cmake官方新增了CUDA的Modules，使得可以使用find_package(CUDA)来获取构建CUDA项目所需要用到的cmake宏，变量以及函数。从cmake版本3.9之后，官方将CUDA加入了cmake的内置语言，从而可以使用project(PROJECT_NAME LANGUAGES CXX CUDA)或者enable_language(CUDA)来让使用CUDA语言。

MUSA沿用了Modules的方式，也提供了相似的cmake模块供使用。保留了和CUDA几乎完全一致的使用方式，以达到用户尽可能方便地迁移CUDA项目的目的。在做项目迁移时，若项目使用cmake工具构建，则绝大多数情况下可以仅做文本替换，将CMakelist.txt中的CUDA替换成MUSA，CU前缀替换成MU前缀。

4.3.1. 使用MUSA模块​

下面从一个最简单的CMakeList.txt入手来了解MUSA模块的使用。

cmake_minimum_required(VERSION 3.10)

set(PROJECT_NAME demo)
project(${PROJECT_NAME} LANGUAGES CXX)

# 载入MUSA模块
list(APPEND CMAKE_MODULE_PATH /usr/local/musa/cmake)
find_package(MUSA REQUIRED)

# 添加可执行文件
musa_add_executable(${PROJECT_NAME} vector_add.mu)

这里假设整个项目仅有一个MUSA代码文件，如上一节使用的vector_add.mu。

  1. 第一步需要使用: find_package(MUSA REQUIRED)，来载入MUSA模块。这里有一点需要注意，由于该模块暂时未被cmake官方收录，仅安装在了MUSA Toolkit的安装目录中，因此需要在载入模块之前将模块的安装目录更新到cmake的MODULE搜索路径中: list(APPEND CMAKE_MODULE_PATH /usr/local/musa/cmake)。
  2. 载入MUSA模块后会添加各种有用的函数，宏以及变量，4.3.2小节将做详细介绍。其中最常用的musa_add_executable，添加目标可执行文件。

4.3.2. MUSA模块提供的物件​

4.3.2.1. 输出变量​

当MUSA模块加载完毕，如下变量将被设置，可供后续使用。

  1. MUSA_FOUND : 若MUSA模块加载成功则为true。
  2. MUSA_VERSION_MAJOR : MUSA的大版本。
  3. MUSA_VERSION_MINOR : MUSA的小版本。
  4. MUSA_VERSION, MUSA_VERSION_STRING : MUSA_VERSION_MAJOR.MUSA_VERSION_MINOR。
  5. MUSA_TOOLKIT_ROOT_DIR : MUSA的安装目录，默认为/usr/local/musa
  6. MUSA_INCLUDE_DIRS : MUSA头文件目录，默认为/usr/local/musa/include
  7. MUSA_LIBRARIES : MUSA运行时库
  8. MUSA_MUFFT_LIBRARIES : muFFT数学库
  9. MUSA_MUBLAS_LIBRARIES : muBLAS数学库，对数学库的使用建议采用MUSAToolkit模块的方式，见4.3.3小节。

4.3.2.2. 函数宏​

cmake中使用函数宏，其名字不区分大小写。

  1. MUSA_ADD_EXECUTABLE : 添加目标可执行文件。

    musa_add_executable(${PROJECT_NAME} file1.mu file2.cu main.cpp)
    # 如上可添加多个编译单元。若编译单元文件后缀名为.mu或者.cu，则会自动被识别成MUSA代码，
    # 也即该编译单元内含有device端运行的kernel代码，从而被mcc编译。对于其他的后缀名如.c,
    # .cc,.cpp等，则仅被当成常规的c代码被host端的c或c++编译器编译。最后这些编译单元的obj产物
    # 会被一起链接成为最终的可执行文件。值得注意的是，这个宏会自动添加MUSA相关的头文件以及
    # 自动链接MUSA运行时库，用户无需额外添加。

    set_source_files_properties(file2.c PROPERTIES MUSA_SOURCE_PROPERTY_FORMAT OBJ)
    musa_add_executable(${PROJECT_NAME} file1.mu file2.c main.cpp)
    # 有时候编译文件虽然后缀名不是.mu或.cu，但是文件里面也写有kernel代码，如上面的file2.c。
    # 这时候需要对该文件做标识，让cmake把此文件当成MUSA代码用mcc来编译。要使用
    # set_source_files_properties(... PROPERTIES MUSA_SOURCE_PROPERTY_FORMAT OBJ) 来标识。
    # 需注意set_source_files_properties要放在musa_add_executable语句之前。

  2. MUSA_ADD_LIBRARY : 添加目标库文件，细节与添加可执行文件相同。

    musa_add_library(${LIB_NAME} SHARED file1.mu file2.cu)
    # SHARED标识编译动态链接库，STATIC标识编译静态链接库。

    set_source_files_properties(file2.c PROPERTIES MUSA_SOURCE_PROPERTY_FORMAT OBJ)
    musa_add_library(${LIB_NAME} SHARED file1.mu file2.c)
    # 指定非.mu 后缀文件为MUSA代码。
    # 需注意set_source_files_properties要放在musa_add_library语句之前。

  3. MUSA_COMPILE : 编译所有输入的MUSA代码，并返回输出obj文件列表，供add_library或者add_executable使用。

    musa_compile(MUSA_OBJ file1.mu file2.cu)
    add_excutable(${PROJECT_NAME} main.cpp ${MUSA_OBJ})
    # file1.mu file2.cu 包含MUSA代码，并提供c++接口。
    # main.cpp 仅调用了上面提供的c++接口。
    # musa_compile将输入的文件用mcc编译成obj文件，
    # 输出obj文件列表返回到第一个参数MUSA_OBJ。
    # 后面的所有c++代码均用host端的c++编译器编译。

  4. MUSA_INCLUDE_DIRECTORIES : 为MUSA代码编译添加头文件目录，将传递给mcc编译器。

    musa_include_directories(${CMAKE_CURRENT_SOURCE_DIR}/internal)
    musa_add_executable(${PROJECT_NAME} main.mu)
    # 假设项目目录结构如下，其中main.mu包含inc.h
    # project
    # |-- main.mu
    # |-- CMakeList.txt
    # |-- internal
    #     |-- inc.h    
    # 需注意musa_include_directories要放在musa_add_executable语句之前。

  5. MUSA_ADD_MUFFT_TO_TARGET : 将muFFT数学库链接到目标。

    musa_add_executable(${PROJECT_NAME} file1.mu main.cpp)
    musa_add_mufft_to_target(${PROJECT_NAME})
    # 等价于添加链接目标-lmufft。 
    # 需注意musa_add_mufft_to_target要放在musa_add_executable语句之后。

  6. MUSA_ADD_MUBLAS_TO_TARGET : 将muBLAS数学库链接到目标。

    musa_add_executable(${PROJECT_NAME} file1.mu main.cpp)
    musa_add_mublas_to_target(${PROJECT_NAME})
    # 等价于添加链接目标-lmublas。
    # 需注意musa_add_mublas_to_target要放在musa_add_executable语句之后。

4.3.2.3. 输入变量​

MUSA模块中提供了一些输入变量，通过这些变量可以控制MUSA_ADD_EXECUTABLE， MUSA_ADD_LIBRARY和MUSA_COMPILE的一些细节行为， 这些变量的设定需要放在那三个函数宏的使用之前。

  1. MUSA_VERBOSE_BUILD : 构建MUSA代码的时候是否打印信息，默认OFF不打印。

    set(MUSA_VERBOSE_BUILD ON)
    musa_add_executable(${PROJECT_NAME} file1.mu main.cpp)
    # 需注意set(MUSA_VERBOSE_BUILD)要放在musa_add_executable语句之前。

  2. MUSA_SOURCE_PROPERTY_FORMAT : 编译文件的格式属性，设置后文件会被强制当成MUSA代码文件。

    set_source_files_properties(file2.c PROPERTIES MUSA_SOURCE_PROPERTY_FORMAT OBJ)
    musa_add_executable(${PROJECT_NAME} file1.mu file2.c main.cpp)
    # 有时候编译文件虽然后缀名不是.mu或.cu，但是文件里面也写有kernel代码，如上面的file2.c。
    # 这时候需要对该文件做标识，让cmake把此文件当成MUSA代码用mcc来编译。要使用
    # set_source_files_properties(... PROPERTIES MUSA_SOURCE_PROPERTY_FORMAT OBJ) 来标识。
    # 需注意set_source_files_properties要放在musa_add_executable语句之前。

    set_source_files_properties(file2.c PROPERTIES MUSA_SOURCE_PROPERTY_FORMAT OBJ)
    musa_add_library(${LIB_NAME} SHARED file1.mu file2.c)
    # 需注意set_source_files_properties要放在musa_add_library语句之前。

  3. MUSA_MCC_FLAGS, MUSA_MCC_FLAGS_CONFIG : 给mcc编译器传入额外的编译选项，CONFIG可以是Debug, MinSizeRel, Release或RelWithDebInfo。默认为空。

    set(MUSA_MCC_FLAGS --offload-arch=mp_21)
    musa_add_executable(${PROJECT_NAME} file1.mu file2.c main.cpp)
    # 针对quyuan1对应的mp_21架构编译。
    # 需注意set(MUSA_MCC_FLAGS)要放在musa_add_executable语句之前。

  4. MUSA_GENERATED_OUTPUT_DIR : 设置mcc编译生成MUSA目标obj文件的存放路径，默认为CMAKE_CURRENT_BINARY_DIR。

  5. MUSA_LINK_LIBRARIES_KEYWORD : 设置cmake链接时的传递属性，可以为PRIVATE，PUBLIC，INTERFACE，默认为空。

    musa_add_library(${LIB_NAME} SHARED file1.mu)
    target_link_libraries(${LIB_NAME} lib)
    # musa_add_library或musa_add_executable执行过程默认链接了MUSA运行时库，
    # 并且默认链接传递属性为空，后面target_link_libraries给目标链接其他库的时候
    # 也需要保持一致，这里不给传递属性。

    set(MUSA_LINK_LIBRARIES_KEYWORD PRIVATE)
    musa_add_library(${LIB_NAME} SHARED file1.mu)
    target_link_libraries(${LIB_NAME} PRIVATE lib)
    # 有时候用户需要设置目标的链接传递属性，比方说PRIVATE, 
    # 使用target_link_libraries(... PRIVATE ...)链接其他库。
    # cmake要求每个目标的所有库的链接传递属性都保持一致，
    # 因此对MUSA运行时库的链接传递属性也需要设置为PRIVATE，
    # 使用set(MUSA_LINK_LIBRARIES_KEYWORD PRIVATE)进行设置。
    # 注意该语句需要放在musa_add_library之前。

4.3.3. 使用MUSAToolkit模块​

cmake版本3.17之后，官方新增加了CUDAToolkit模块。其用途是，有些工程并没有包含任何device端的kernel代码，但是调用了NVIDIA官方提供的数学库或图像处理库等现成的c或c++接口，从而使用GPU进行加速。这种情况下整个工程内全部代码都是c或c++语言编写，无任何CUDA代码，故无需使用nvcc编译器编译工程。那么整个项目仅需要使用host端的c或c++编译器编译，最后链接的时候把运行时库以及数学库等添加上即可。CUDAToolkit模块则提供了所有使用GPU加速可能使用到的库目标。

为了保证对CUDA使用的最佳兼容，MUSAToolkit中也包含了cmake的MUSAToolkit模块。基本使用方式如下:

cmake_minimum_required(VERSION 3.10)

set(PROJECT_NAME demo)
project(${PROJECT_NAME} LANGUAGES CXX)

# 载入MUSAToolkit模块
list(APPEND CMAKE_MODULE_PATH /usr/local/musa/cmake)
find_package(MUSAToolkit REQUIRED)

# 添加可执行文件
add_executable(${PROJECT_NAME} main.cpp)

# 为目标链接运行时库及数学库
target_link_libraries(${PROJECT_NAME} 
  MUSA::musart
  MUSA::mufft
  )

示例项目中调用了傅里叶变换库进行加速。

  1. 第一步需要使用: find_package(MUSAToolkit REQUIRED)，来载入MUSAToolkit模块。同样的由于该模块暂时未被cmake官方收录，仅安装在了MUSA Toolkit的安装目录中，因此需要在载入模块之前将模块的安装目录更新到cmake的MODULE搜索路径中: list(APPEND CMAKE_MODULE_PATH /usr/local/musa/cmake)。
  2. 模块载入之后，将提供若干库目标以及变量供使用，这个例子中用到了运行时库和傅里叶变换库，故给目标添加链接库MUSA::musart和MUSA::mufft。需要留意，这里模块提供的目标MUSA::已经包含了所需的头文件路径，会自动传递给要编译的目标，故无需再给编译目标添加MUSA相关的头文件目录。

4.3.4. MUSAToolkit模块提供的物件​

和MUSA模块不同，MUSAToolkit模块没有提供复杂的函数宏，而是仅提供了若干输出变量以及库目标。

4.3.4.1. 输出变量​

当MUSAToolkit模块加载完毕，如下变量将被设置，可供后续使用。

  1. MUSAToolkit_FOUND : 若MUSAToolkit_FOUND模块加载成功则为true。
  2. MUSAToolkit_VERSION_MAJOR : MUSAToolkit的大版本。
  3. MUSAToolkit_VERSION_MINOR : MUSAToolkit的小版本。
  4. MUSAToolkit_VERSION_PATCH : MUSAToolkit的patch版本。
  5. MUSAToolkit_VERSION : _MAJOR._MINOR._PATCH。
  6. MUSAToolkit_BIN_DIR : MUSAToolkit工具链可执行文件目录，默认为/usr/local/musa/bin。
  7. MUSAToolkit_INCLUDE_DIRS : MUSAToolkit头文件目录，默认为/usr/local/musa/bin/include。
  8. MUSAToolkit_LIBRARY_DIR : MUSAToolkit链接库存放目录，默认为/usr/local/musa/bin/lib。
  9. MUSAToolkit_TARGET_DIR : MUSAToolkit的根目录，默认为/usr/local/musa。
  10. MUSAToolkit_MCC_EXECUTABLE : MUSA编译器文件。

4.3.4.2. 库目标​

使用target_link_libraries为目标添加如下需要的库。

  1. MUSA::musart : MUSA运行时库，包含如musaMalloc，musaFree等api。
  2. MUSA::musa_driver : MUSA驱动库，包含如muMemAlloc，muMemFree等api。
  3. MUSA::mublas : 线性代数库。
  4. MUSA::mufft : 傅里叶变换库。
  5. MUSA::murand : 随机数生成库。
  6. MUSA::musolver : 求解器相关库。
  7. MUSA::musparse : 稀疏矩阵相关库。
  8. MUSA::mupti : 性能分析库。
  9. MUSA::muppc, MUSA::muppial, MUSA::muppicc, MUSA::muppicom, MUSA::muppidei, MUSA::muppif, MUSA::muppig, MUSA::muppim, MUSA::muppist, MUSA::muppisu, MUSA::muppitc, MUSA::mupps : 二维图像及信号处理。

版权所有 2024 摩尔线程，本文档受国际版权法保护。

摩尔线程和摩尔线程徽标是摩尔线程智能科技 (北京)有限责任公司的注册商标。

免责声明

本文档提供有关摩尔线程产品的信息。本文档并未授权任何知识产权的许可，并未以明示或暗示，或以禁止反言或其他方式授予任何知识产权许可。