如何将内核卸载到 GPU 来调试以下 DPC++ 程序？

Question

我想运行英特尔 GNU 项目调试器中的 DPC++ 程序。我已经从 Intel OneAPI Basekit 下载了 Intel GDB。它预装了 OneAPI Basekit。

要下载的link是：

[https://software.intel.com/content/www/us/en/develop/tools/oneapi/base-toolkit.html#gs.ynm6aj]

如何在内核卸载到GPU的情况下调试下面的DPC++程序？如何在下级和线程之间切换？请查看下面提供的代码。

#include <CL/sycl.hpp>
#include <iostream>
// Location of file: <oneapi-root>/dev-utilities/<version>/include
#include "dpc_common.hpp"
#include "selector.hpp"

using namespace std;
using namespace sycl;

// A device function, called from inside the kernel.
static size_t GetDim(id<1> wi, int dim) {
  return wi[dim];
}

int main(int argc, char *argv[]) {
  constexpr size_t length = 64;
  int input[length];
  int output[length];

  // Initialize the input
  for (int i = 0; i < length; i++)
    input[i] = i + 100;

  try {
    CustomSelector selector(GetDeviceType(argc, argv));
    queue q(selector, dpc_common::exception_handler);
    cout << "[SYCL] Using device: ["
         << q.get_device().get_info<info::device::name>()
         << "] from ["
         << q.get_device().get_platform().get_info<info::platform::name>()
         << "]\n";

    range data_range{length};
    buffer buffer_in{input, data_range};
    buffer buffer_out{output, data_range};

    q.submit([&](auto &h) {
      accessor in(buffer_in, h, read_only);
      accessor out(buffer_out, h, write_only);

      // kernel-start
      h.parallel_for(data_range, [=](id<1> index) {
        size_t id0 = GetDim(index, 0);
        int element = in[index];  // breakpoint-here
        int result = element + 50;
        if (id0 % 2 == 0) {
          result = result + 50;  // then-branch
        } else {
          result = -1;  // else-branch
        }
        out[index] = result;
      });
      // kernel-end
    });

    q.wait_and_throw();
  } catch (sycl::exception const& e) {
    cout << "fail; synchronous exception occurred: " << e.what() << "\n";
    return -1;
  }

  // Verify the output
  for (int i = 0; i < length; i++) {
    int result = (i % 2 == 0) ? (input[i] + 100) : -1;
    if (output[i] != result) {
      cout << "fail; element " << i << " is " << output[i] << "\n";
      return -1;
    }
  }

  cout << "success; result is correct.\n";
  return 0;
}

Answer 1

可以使用调试器列出应用程序的线程。打印的信息包括线程 ID 和线程当前停止的位置。对于 GPU 线程，调试器还会打印活动的 SIMD 通道。 GDB 使用以下格式显示线程：

<inferior_number>.<thread_number>:<SIMD Lane/s>

您可以切换线程以及 SIMD 通道以使用“thread”命令（例如“thread 3:4”、“thread :6”或“thread 7”）更改上下文。第一个命令切换到线程 3 和 SIMD 通道 4。第二个命令切换到当前线程中的 SIMD 通道 6。第三个命令切换到线程 7。选择的默认通道将是先前选择的通道（如果它处于活动状态），或者是线程中的第一个活动通道。更多详细信息，请参阅下面的link，其中解释了如何在 GPU 设备上调试代码。

[https://software.intel.com/content/www/us/en/develop/documentation/debugging-dpcpp-linux/top/debug-a-dpc-application-on-a-gpu/basic-debugging-1.html]

如何将内核卸载到 GPU 来调试以下 DPC++ 程序？

How to debug the below DPC++ program with the kernel offloaded to the GPU?

gdb

kernel

intel-oneapi

dpc++

intel-gdb