如何在NVRTC编译的程序中正确使用include stdio. h？

ttvkxqim 于 2023-08-03 发布在其他

关注(0)|答案(2)|浏览(103)

我已经写了一个惊人的内核，这将给我带来名誉和财富-如果我只能用NVRTC编译它：

#include <stdio.h>

__global__ void do_stuff() { }

字符串
我本来希望系统头可以被（运行时）编译器识别，就像一个普通的编译器一样，并且这将“正常工作”（模任何printf特定的机器）。或者，如果它不起作用，我会期望得到一个错误消息，说明stdio.h的源代码不能通过“程序创建”API调用（nvrtcCreateProgram()）获得，因为我将NULL和NULL作为它的最后两个参数传递。
然而，我得到的是以下内容：

/usr/include/stdio.h(33): catastrophic error: cannot open source file "stddef.h"

型
我觉得很奇怪。这意味着运行时编译器 * 能够 * 查看系统头文件，但 * 不能 * 找到stddef.h，就像nvcc或主机端编译器一样。
为什么会发生这种情况，惯用/推荐的解决方法是什么？
注意：我想一个解决办法，这将是跨平台，而不仅仅是我的个人机器上工作。

c

来源：https://stackoverflow.com/questions/62860627/how-to-properly-use-include-stdio-h-in-an-nvrtc-compiled-program

2条答案

按热度按时间

m1m5dgzv1#

Robert Crovella亲切地提醒我的“JITify”库中采用了另一种方法。虽然这似乎没有很好地记录，但Jitify预先包含了它认为适合的各种标题的处理片段。特别是对于<climits>/<limits.h>：

static const char* jitsafe_header_limits_h = R"(
#pragma once
#if defined _WIN32 || defined _WIN64
 #define __WORDSIZE 32
#else
 #if defined __x86_64__ && !defined __ILP32__
  #define __WORDSIZE 64
 #else
  #define __WORDSIZE 32
 #endif
#endif
#define MB_LEN_MAX  16
#define CHAR_BIT    8
#define SCHAR_MIN   (-128)
#define SCHAR_MAX   127
#define UCHAR_MAX   255
enum {
  _JITIFY_CHAR_IS_UNSIGNED = (char)-1 >= 0,
  CHAR_MIN = _JITIFY_CHAR_IS_UNSIGNED ? 0 : SCHAR_MIN,
  CHAR_MAX = _JITIFY_CHAR_IS_UNSIGNED ? UCHAR_MAX : SCHAR_MAX,
};
#define SHRT_MIN    (-32768)
#define SHRT_MAX    32767
#define USHRT_MAX   65535
#define INT_MIN     (-INT_MAX - 1)
#define INT_MAX     2147483647
#define UINT_MAX    4294967295U
#if __WORDSIZE == 64
 # define LONG_MAX  9223372036854775807L
#else
 # define LONG_MAX  2147483647L
#endif
#define LONG_MIN    (-LONG_MAX - 1L)
#if __WORDSIZE == 64
 #define ULONG_MAX  18446744073709551615UL
#else
 #define ULONG_MAX  4294967295UL
#endif
#define LLONG_MAX  9223372036854775807LL
#define LLONG_MIN  (-LLONG_MAX - 1LL)
#define ULLONG_MAX 18446744073709551615ULL
)";

字符串
对于stddef.h：

static const char* jitsafe_header_stddef_h =
    "#pragma once\n"
    "#include <climits>\n"
    "namespace __jitify_stddef_ns {\n"
    "#if __cplusplus >= 201103L\n"
    "typedef decltype(nullptr) nullptr_t;\n"
    "#if defined(_MSC_VER)\n"
    "  typedef double max_align_t;\n"
    "#elif defined(__APPLE__)\n"
    "  typedef long double max_align_t;\n"
    "#else\n"
    "  // Define max_align_t to match the GCC definition.\n"
    "  typedef struct {\n"
    "    long long __jitify_max_align_nonce1\n"
    "        __attribute__((__aligned__(__alignof__(long long))));\n"
    "    long double __jitify_max_align_nonce2\n"
    "        __attribute__((__aligned__(__alignof__(long double))));\n"
    "  } max_align_t;\n"
    "#endif\n"
    "#endif  // __cplusplus >= 201103L\n"
    "#if __cplusplus >= 201703L\n"
    "enum class byte : unsigned char {};\n"
    "#endif  // __cplusplus >= 201703L\n"
    "} // namespace __jitify_stddef_ns\n"
    "namespace std {\n"
    "  // NVRTC provides built-in definitions of ::size_t and ::ptrdiff_t.\n"
    "  using ::size_t;\n"
    "  using ::ptrdiff_t;\n"
    "  using namespace __jitify_stddef_ns;\n"
    "} // namespace std\n"
    "using namespace __jitify_stddef_ns;\n";

型
对于stdio.h：

static const char* jitsafe_header_stdio_h =
    "#pragma once\n"
    "#include <stddef.h>\n"
    "#define FILE int\n"
    "int fflush ( FILE * stream );\n"
    "int fprintf ( FILE * stream, const char * format, ... );\n";

型
如果您包含这些字符串作为头文件，并使用适当的名称作为键，那么您的内核很可能会编译。
事实上，可以在jitify.hpp中使用这些和其他迷你头文件形成头文件，用于非NVRTC内核编译。这也可能有用。
最后一点：上面的常量不指定__device__执行空间。因此，要么在其中添加__device__，要么告诉编译器假设函数仅在设备上执行，除非另有说明;这是--device-as-default-execution-space NVRTC编译器选项。

赞(0）回复(0）举报 2023-08-03

wfsdck302#

这里有两个可能有效的解决方案，但我宁愿“避免”。如果他们是唯一合理的行动方案毕竟-请评论并说：
1.添加stddef.h的特定路径作为编译器参数（-I或--include-path=）。
1.将stddef.h的源传递给nvrtcCreateProgram()调用。