数据复制

复制方向

常见复制方向如下：

同步接口aclrtMemcpy在复制完成后返回。异步接口aclrtMemcpyAsync将复制任务下发至指定的Stream，接口返回仅表示任务已下发，并不表示复制任务已完成；在读取目标内存前，需要调用aclrtSynchronizeStream、aclrtSynchronizeDevice或通过Event等机制来确认任务已完成。

Host与Device复制

Host与Device之间的数据复制是最常见的输入、输出搬运方式。典型流程为：申请Host内存和Device内存，将输入从Host复制到Device，在Stream上下发算子，最后将输出从Device复制回Host。

aclrtStream stream = nullptr;
aclrtCreateStream(&stream);

void *hostIn = nullptr;
void *hostOut = nullptr;
void *devIn = nullptr;
void *devOut = nullptr;
aclrtMallocHost(&hostIn, size);
aclrtMallocHost(&hostOut, size);
aclrtMalloc(&devIn, size, ACL_MEM_MALLOC_HUGE_FIRST);
aclrtMalloc(&devOut, size, ACL_MEM_MALLOC_HUGE_FIRST);

// 输入数据从Host复制到Device
aclrtMemcpyAsync(devIn, size, hostIn, size, ACL_MEMCPY_HOST_TO_DEVICE, stream);

// 下发计算任务，和前序复制任务在同一Stream中保序执行
myKernel<<<numBlocks, nullptr, stream>>>(devIn, devOut);

// 输出数据从Device复制回Host
aclrtMemcpyAsync(hostOut, size, devOut, size, ACL_MEMCPY_DEVICE_TO_HOST, stream);

// 等待同一Stream上复制和计算全部完成，再读取hostOut
aclrtSynchronizeStream(stream);

若使用同步复制接口，接口会在本次复制完成后返回，但不会对其他Stream进行隐式同步。若复制前后还有其他异步任务，需要通过Stream顺序、Event或显式同步来确保数据依赖关系。

非锁页Host内存的异步复制

通过malloc、mmap等传统接口申请的Host内存是非锁页内存。aclrtMemcpyAsync、aclrtMemcpy2dAsync、aclrtMemcpyBatchAsync等异步接口支持非锁页Host内存，但这类场景需要单独关注：

• 当参与复制的Host内存是锁页内存时，异步复制接口只下发复制任务并返回，Host线程可以继续执行其它工作，复制任务在Stream中按序完成。
• 当参与复制的Host内存是非锁页内存时，异步复制接口会在内存复制任务完成后才返回。此时接口名虽然带Async，但Host线程无法与本次复制并行执行，数据传输也难以与计算充分重叠。
• 若需要稳定获得复制与计算重叠效果，建议通过aclrtMallocHost申请锁页内存，或在支持的系统上使用aclrtHostRegisterV2将已有Host内存注册为锁页内存。
• ACL Graph捕获异步内存复制任务时，若复制涉及Host内存，Host内存需要使用Runtime接口申请的锁页内存，否则捕获过程中会返回错误。

因此，非锁页Host内存更适合低频、临时、对并发性能不敏感的数据搬运；性能敏感的输入输出通路应使用锁页Host内存。

Device内复制

使用aclrtMemcpy、aclrtMemcpyAsync进行Device到Device复制时，复制类型使用ACL_MEMCPY_DEVICE_TO_DEVICE即可。该枚举表示同Device内或跨Device的数据复制，Runtime会根据源地址和目的地址判断实际复制路径，用户不需要显式指定ACL_MEMCPY_INNER_DEVICE_TO_DEVICE或ACL_MEMCPY_INTER_DEVICE_TO_DEVICE。

ACL_MEMCPY_INNER_DEVICE_TO_DEVICE只表示同Device内复制，需要在描述符复制场景中显式指定，例如aclrtMemcpyAsyncWithDesc接口。描述符方式适合复制参数可复用的场景。用户先获取描述符大小并设置源地址、目的地址和长度，再在Stream上使用描述符下发复制任务。

跨Device复制

跨Device复制用于同一进程内不同Device之间的数据搬运。调用复制接口前，应先确认两个Device之间是否支持数据交互，并按访问方向开启两个Device之间的数据交互：

int32_t canAccess = 0;
aclrtDeviceCanAccessPeer(&canAccess, srcDeviceId, dstDeviceId);
if (canAccess != 0) {
    aclrtSetDevice(dstDeviceId);
    aclrtDeviceEnablePeerAccess(srcDeviceId, 0);
}

跨Device内存复制的详细描述请参见跨Device的数据交互。

二维和批量复制

二维复制接口aclrtMemcpy2d、aclrtMemcpy2dAsync适用于矩阵、图像或带pitch的二维数据。调用时需要分别指定源和目的内存中相邻两列向量的地址距离，以及待复制区域的宽度和高度。异步二维复制涉及Host内存时，同样遵循“锁页Host内存才是真正异步”的规则。

批量复制接口aclrtMemcpyBatch、aclrtMemcpyBatchAsync及其V2版本适用于一次下发多段Host到Device或Device到Host复制。批量复制需要注意：

• 批次内的复制操作不保证按照数组顺序执行。
• 每个复制操作的目的地址、源地址和复制大小数组长度必须一致。
• 同一个批次中的复制方向仅支持Host到Device或Device到Host中的一种。
• 相关接口属于试验特性时，应以API参考中的产品说明和约束为准。

内存初始化

aclrtMemset、aclrtMemsetAsync用于初始化Device内存。同步接口在初始化完成后返回；异步接口把内存初始化任务下发到指定Stream，接口返回仅表示任务已下发，并不表示复制任务已完成。

对于作为同步标记或计算输入的Device内存，建议在使用前显式初始化，避免读取到未定义数据。