CUDA统一内存编程 - 我可以将指针传递给不了解CUDA的模块吗？

Question

我将分配一些内存cudaMallocManaged.我可以安全地将此指针传递给另一个程序模块(编译成另一个.o文件),该程序模块不知道CUDA,只会使用简单的memcpy或其他东西来操纵指针后面的数据吗？

这样的事情

// compiled into A.o
class A{
  void* getMem(int size){
    void* ptr;
    cudaMallocManaged(*ptr, size);
    return ptr;
  }

  // some kernels here

}


// compiled into B.o
class B{
  void manipulateMem(void* ptr, void* source, int size){
    memcpy(ptr, source, size);
  }

}

然后像这样的代码,可能编译成main.o:

A a;
B b;
void* mem = a.getMem(100);
b.manipulateMem(mem, source, 100);

我没有发现自动复制/同步在这种情况下不起作用的任何通知.

Answer 1

我可以安全地将此指针传递给另一个程序模块(编译成另一个.o文件),该程序模块不知道CUDA,只会使用简单的memcpy或其他东西来操纵指针后面的数据吗？

是的,您可以,但仍必须遵守统一内存访问规则(< - 提示:单击此处并阅读此内容).对于计算能力为3.0或更高但低于6.0的GPU,这些规则目前是:

1. 最初,在分配(使用诸如托管分配器cudaMallocManaged)之后,在调用任何内核之前,可以从主机CPU上运行的代码访问托管指针.
2. 一旦调用了任何内核,并且直到调用了后续内容cudaDeviceSynchronize(),主机代码就无法访问数据,并且在主机代码中使用它的任何尝试都将导致UB,其可能包括seg错误.
3. 在同步事件(例如cudaDeviceSynchronize())之后,恢复对受管指针引用的数据的主机访问.(从技术上讲,在当前的实现中,此时主机对数据的访问通常会导致页面错误,这是一个操作系统可见的事件.这些页面错误基本上调用了CUDA运行时,后者cudaMemcpy在引擎盖下执行必要的操作,将数据返回给主机,以便为页面错误提供服务.但是在括号中的这些注释对于理解行为的一般规则不是必需的.)

因此,CUDA运行时基本上具有显式标记(内核启动 - >同步),它明确地标识了如何管理数据(何时迁移以及在哪个方向上).因此,即使您在某些"模块中运行代码而不知道CUDA",如果遵守上述规则,该模块中的代码也可以访问数据,因为CUDA运行时具有足够的,明确的标识如何管理数据的标记.

对于计算能力为6.0或更高的GPU,上述规则基本上不适用.对于这些GPU,主机和设备的并发访问是可能的,但是对公共资源的多次访问仍然存在竞争条件的可能性,如在任何多处理器/多线程环境中那样. 目前,CUDA运行时不强制主机和设备访问同一内存页面之间的任何特定访问排序规则.