1

假设我采用了一个 CUDA 程序——例如 CUDAvectorAdd示例，并删除了内核的实现，但仍然有启动命令：

vectorAdd<<<blocksPerGrid, threadsPerBlock>>>(d_A, d_B, d_C, numElements);

并假设我编写自己的 PTX，因为我是一个 DIY 类型的人，所以现在我vectorAdd.cu没有内核的 CUDA 代码和vectorAdd.ptx.

我现在可以生成一个可执行文件，它可以像未修改的 vectorAdd 那样工作，但在 PTX 中运行代码吗？

（假设 PTX 没有尝试任何有趣的事情或做任何错误的讨论，同样的讨论。）

笔记：

  • 这个问题是关于：

    如何创建可执行文件以在给定的 PTX 文件中运行内核？

    除此之外，在那个问题中，发帖人愿意使用驱动程序 API 来动态加载和编译使用驱动程序 API 的 PTX 文件。在这里，这不是一个选项：C++ 代码使用三人字形 CUDA 运行时启动，这不能改变。

  • 我不介意创建涉及生成其他文件（例如 cubin）的可执行文件的过程。

4

1 回答 1

1
  1. 将您的功能定义为
    __ global __ void vectorAdd(void* d_A, void* d_B, void* d_C, int numElements);
    在调用者可见的标题中
  2. 创建一个带有空声明的文件 vectorAdd.cu
    __ global __ void vectorAdd(void* d_A, void* d_B, void* d_C, int numElements) {}
  3. 称呼
    nvcc --keep vectorAdd.cu
    有合适的选择
  4. 将 vectorAdd.ptx 替换为您的版本
  5. 称呼
    nvcc -fatbin -dlink
    创建 fatbin 和 cubin 文件
  6. 调用 nvcc -link 链接 .cubin 文件和 .cudafe1.cpp 或 cudafe1.c（取决于语言）文件。它们还依次包含 .cudafe1.stub.c 和 .fatbin.c 文件
  7. 在项目中使用生成的 .obj 或 .o 文件 (Windows/Linux)
  8. 以CUDA运行时方式调用vectorAdd<<<>>>

（作为高级 DIY 人员，您将来会想要编写 SASS 代码，这是特定于设备的低级汇编语言。）

于 2021-06-03T00:05:56.513 回答