跳到主要内容

musify 一键代码迁移

什么是 musify?​

musify 是摩尔线程推出的一款 CUDA 到 MUSA 一键代码迁移工具,属于 MUSA Toolkits 的一部分。musify 通过命令行工具 musify-text 提供服务,采用纯文本匹配的方式进行代码转换,实现:

  • 轻量级: 避免语法分析引入过重的依赖
  • 易部署: 无需复杂的编译器基础设施
  • 灵活快速: 工具小巧灵活
注意

由于仅基于文本匹配,musify 无法理解代码语义,可能会误改不该转换的内容。你可以在代码中添加排除标记(exclude markers),明确告诉工具哪些部分跳过处理。


主要特性​

  • 自动转换 - 自动将 CUDA API 调用转换为 MUSA API
  • 语法兼容 - 支持 CUDA C/C++ 语法转换
  • 批量处理 - 支持批量转换多个源文件
  • 原地转换 - 支持 --inplace 模式直接修改源文件
  • 双向转换 - 支持 CUDA→MUSA 和 MUSA→CUDA 双向转换
  • 自定义映射 - 支持用户自定义 API 映射表

快速开始​

安装​

安装 MUSA SDK 后,musify-text 命令即可使用。

提示

如果需要在其他机器上单独使用 musify,或者 SDK 安装有问题,可以手动安装依赖。

(可选)手动安装依赖​

musify 使用 Python 编写,需要系统中安装有 Python 3。

# Ubuntu 系统安装依赖
sudo apt install python-is-python3 -y
sudo apt install pip -y
pip install ahocorapy

# 如遇网络问题,可配置国内镜像源
pip config set global.index-url https://pypi.mirrors.ustc.edu.cn/simple/

基本示例​

# 转换单个文件(输出到新文件,默认行为)
musify-text source.cu

# 原地转换(直接修改源文件)
musify-text --inplace source.cu

# 输出到标准输出
musify-text --terminal source.cu

# 转换多个文件
musify-text --inplace -- *.cu

转换示例​

CUDA 源代码 (source.cu):

#include <cuda_runtime.h>

__global__ void vector_add(float* a, float* b, float* c, int n) {
int idx = blockIdx.x * blockDim.x + threadIdx.x;
if (idx < n) {
c[idx] = a[idx] + b[idx];
}
}

int main() {
float *d_a, *d_b, *d_c;
cudaMalloc(&d_a, N * sizeof(float));
cudaMalloc(&d_b, N * sizeof(float));
cudaMalloc(&d_c, N * sizeof(float));

vector_add<<<grid, block>>>(d_a, d_b, d_c, N);

cudaFree(d_a);
cudaFree(d_b);
cudaFree(d_c);

return 0;
}

转换后 MUSA 代码:

#include <musa_runtime.h>

__global__ void vector_add(float* a, float* b, float* c, int n) {
int idx = blockIdx.x * blockDim.x + threadIdx.x;
if (idx < n) {
c[idx] = a[idx] + b[idx];
}
}

int main() {
float *d_a, *d_b, *d_c;
musaMalloc((void**)&d_a, N * sizeof(float));
musaMalloc((void**)&d_b, N * sizeof(float));
musaMalloc((void**)&d_c, N * sizeof(float));

vector_add<<<grid, block>>>(d_a, d_b, d_c, N);

musaFree(d_a);
musaFree(d_b);
musaFree(d_c);

return 0;
}

编译转换后的示例​

# 转换
musify-text --inplace source.cu

# 重命名为 .mu 后缀
mv source.cu source.mu

# 编译
mcc source.mu -lmusart -L/usr/local/musa/lib -o app

命令行选项​

帮助信息​

$ musify-text -h
usage: musify-text [-h] [-t | -c | -i] [-d {c2m,m2c}] [-m [MAPPING ...]]
[--clear-mapping] [-l {DEBUG,INFO,WARNING}] [srcs ...]

positional arguments:
srcs source files to be transformed

options:
-h, --help show this help message and exit
-t, --terminal print code to stdout
-c, --create write code to newly created file, default action
-i, --inplace modify code inplace
-d {c2m,m2c}, --direction {c2m,m2c}
convert direction
-m [MAPPING ...], --mapping [MAPPING ...]
api mapping
--clear-mapping clear default and previous mapping
-l {DEBUG,INFO,WARNING}, --log-level {DEBUG,INFO,WARNING}
lowest log level to display

选项详解​

选项说明
-t, --terminal输出到标准输出(stdout)
-c, --create写入新文件(默认行为)
-i, --inplace原地修改源文件
-d, --direction转换方向:c2m(CUDA→MUSA)或 m2c(MUSA→CUDA)
-m, --mapping指定映射表 JSON 文件(可多个)
--clear-mapping清除默认映射表,只用 -m 指定的
-l, --log-level日志级别(DEBUG/INFO/WARNING)
重要提示
  1. 输出方式三选一:-t、-c、-i 必须选择其一
  2. 文件路径保护:在所有选项之后、文件路径之前,加上 --,防止以 - 开头的文件路径被识别为选项
  3. 映射表叠加:-m 会添加新的映射表,不会覆盖默认映射表
  4. 映射表格式:JSON 文件,内容为单层无嵌套的 JSON Object,每个 name-value 对分别表示相应的 CUDA 和 MUSA 命名
{
"cudaMalloc": "musaMalloc",
"cudaFree": "musaFree",
"cudaMemcpy": "musaMemcpy"
}

高级用法​

批量迁移​

使用 find 命令​

# 递归查找目录 ${DIR} 下所有后缀名为 cu、cuh、cpp 或 h 的文件并转化
musify-text --inplace -- `find ${DIR} \
-name '*.cu' \
-o -name '*.cuh' \
-o -name '*.cpp' \
-o -name '*.h'`

(推荐)使用 ripgrep​

ripgrep 是一个现代化的快速文本搜索工具,可以用于文件遍历:

# 安装 ripgrep
sudo apt install ripgrep -y
# 方式 1: 使用 rg --files 配合预设类型(cpp 类型已内置)
musify-text --inplace -- `rg --files -tcpp ${DIR}`

# 方式 2: 使用 rg --files 配合自定义类型(cuda 类型需要手动添加)
musify-text --inplace -- `rg --files \
--type-add 'cuda:*.cu' \
--type-add 'cuda:*.cuh' \
-tcuda -tcpp ${DIR}`

# 方式 3: 使用 -g 选项直接指定后缀名(类似 find)
musify-text --inplace -- `rg --files \
-g '*.cu' \
-g '*.cuh' \
-g '*.cpp' \
-g '*.h' ${DIR}`

自定义映射表​

# 使用自定义映射表(默认映射表仍生效)
musify-text --inplace -m my_mapping.json -- source.cu

# 清除默认映射表,只用自定义映射表
musify-text --inplace --clear-mapping -m my_mapping.json -- source.cu

# 使用多个映射表
musify-text --inplace -m mapping1.json -m mapping2.json -- source.cu

调整日志级别​

# 显示 DEBUG 级别日志(最详细)
musify-text -l DEBUG --inplace -- source.cu

# 只显示 WARNING 及以上级别(最简洁)
musify-text -l WARNING --inplace -- source.cu

排除标记​

由于 musify 是纯文本匹配,可能转换一些不需要转换的代码。为此,我们参考 lcov 行覆盖率工具的设计,加入了排除标记功能。

标记类型​

标记说明
MUSIFY_EXCL_LINE包含本标记的行会被排除
MUSIFY_EXCL_START从包含本标记的行开始,直到 MUSIFY_EXCL_STOP 之间的行会被排除
MUSIFY_EXCL_STOP结束由 MUSIFY_EXCL_START 开启的排除

使用标记​

方式 1:(单独使用)LINE 标记​

// 被排除,不会转换
char str[] = "cuInit"; // MUSIFY_EXCL_LINE

// 被排除,不会转换
printf("This is a cudaMalloc example"); // MUSIFY_EXCL_LINE

方式 2:(成对使用)START/STOP 标记​

// MUSIFY_EXCL_START
char *str_array[] = {
"cuInit",
"cuMalloc",
"cuMemset",
"cuLaunchKernel"
};
// MUSIFY_EXCL_STOP

// 上述代码块中的所有 CUDA API 名称都不会被转换
注意
  1. START 和 STOP 必须成对使用,否则从 START 到文件末尾的所有内容都会被排除
  2. START 和 STOP 所在行也会被排除
  3. LINE 标记所在行会被排除

转换方向​

CUDA → MUSA(默认)​

# 默认方向,无需指定
musify-text --inplace -- source.cu

# 显式指定
musify-text -d c2m --inplace -- source.cu

MUSA → CUDA​

# 反向转换(用于测试或回滚)
musify-text -d m2c --inplace -- source_musa.cpp

API 映射对照表​

常用 API​

CUDA APIMUSA API
cudaMallocmusaMalloc
cudaMallocHostmusaMallocHost
cudaFreemusaFree
cudaFreeHostmusaFreeHost
cudaMallocManagedmusaMallocManaged
cudaMemcpymusaMemcpy
cudaMemcpyAsyncmusaMemcpyAsync
cudaMemcpyToSymbolmusaMemcpyToSymbol
cudaMemcpyFromSymbolmusaMemcpyFromSymbol
cudaMemsetmusaMemset
cudaMemsetAsyncmusaMemsetAsync

设备管理 API​

CUDA APIMUSA API
cudaGetDevicemusaGetDevice
cudaSetDevicemusaSetDevice
cudaGetDeviceCountmusaGetDeviceCount
cudaDeviceGetAttributemusaDeviceGetAttribute

流管理 API​

CUDA APIMUSA API
cudaStreamCreatemusaStreamCreate
cudaStreamDestroymusaStreamDestroy
cudaStreamSynchronizemusaStreamSynchronize
cudaStreamQuerymusaStreamQuery

事件管理 API​

CUDA APIMUSA API
cudaEventCreatemusaEventCreate
cudaEventDestroymusaEventDestroy
cudaEventRecordmusaEventRecord
cudaEventSynchronizemusaEventSynchronize
cudaEventElapsedTimemusaEventElapsedTime

错误处理 API​

CUDA APIMUSA API
cudaGetErrorStringmusaGetErrorString
cudaGetLastErrormusaGetLastError

内存拷贝类型枚举​

CUDAMUSA
cudaMemcpyHostToHostmusaMemcpyHostToHost
cudaMemcpyHostToDevicemusaMemcpyHostToDevice
cudaMemcpyDeviceToHostmusaMemcpyDeviceToHost
cudaMemcpyDeviceToDevicemusaMemcpyDeviceToDevice
cudaMemcpyDefaultmusaMemcpyDefault

最佳实践​

版本控制​

在使用 musify 之前,建议先将代码提交到版本控制系统:

# 使用 git 管理
git add .
git commit -m "Before musify migration"

# 然后再执行迁移
musify-text --inplace -- `rg --files -tcpp src/`

# 如果有问题,可以回滚
git checkout .

增量迁移​

对于大型项目,建议分批次迁移:

# 先迁移一个模块
musify-text --inplace -- `rg --files -tcpp src/module_a/`

# 编译测试
make clean && make

# 确认无误后,继续下一个模块
musify-text --inplace -- `rg --files -tcpp src/module_b/`

排除特殊代码​

对于以下类型的代码,建议使用排除标记:

  • 字符串字面量中的 CUDA API 名称
  • 注释中的 CUDA API 名称
  • 日志输出中的 CUDA API 名称
  • 测试代码中需要保留的 CUDA 引用
// 日志中的 CUDA 引用,建议排除
printf("Original implementation used cuInit API"); // MUSIFY_EXCL_LINE

局限性​

文本匹配的局限​

由于 musify 采用纯文本匹配,存在以下局限:

局限说明解决方案
无法区分上下文字符串、注释中的 API 名也会被转换使用排除标记
无法理解语义可能转换不该转换的标识符转换后人工审查
不支持复杂模式无法处理需要语法分析的场景手动修改

故障排除​

常见问题​

Q1: 转换后代码无法编译

A: 检查是否有需要排除的代码未排除,或存在映射表中未覆盖的 API。可以查看 DEBUG 日志了解具体转换了哪些内容:

musify-text -l DEBUG --inplace -- source.cu 2>&1 | grep -i "converted"
Q2: 某些 CUDA API 没有被转换

A: 检查映射表中是否包含该 API。如需添加,可以创建自定义映射表:

cat >> my_mapping.json << EOF
{
"cudaCustomAPI": "musaCustomAPI"
}
EOF

musify-text -m my_mapping.json --inplace -- source.cu
Q3: 转换引入了错误

A: 使用 git diff 查看具体变更,定位问题后可以:

# 回滚
git checkout source.cu

# 添加排除标记后重新转换
# 编辑 source.cu,在需要排除的行添加 // MUSIFY_EXCL_LINE
musify-text --inplace -- source.cu

日志分析​

# 启用 DEBUG 日志
musify-text -l DEBUG --inplace -- source.cu

# 日志输出示例:
# [INFO] Processing file: source.cu
# [DEBUG] Line 42: cudaMalloc -> musaMalloc
# [DEBUG] Line 43: cudaMemcpy -> musaMemcpy
# [INFO] Converted 23 API calls

相关工具​

MUSA for VS Code 是摩尔线程提供的 Visual Studio Code 插件,支持智能代码编辑、实时错误诊断、以及 CUDA → MUSA 一键迁移功能,可在 IDE 内直接使用 musify 进行代码转换,适合需要频繁进行迁移工作的开发者。


相关文档​