Как правильно использовать include stdio.h в программе, скомпилированной с помощью NVRT C?

Question

Я написал потрясающее ядро, которое принесет мне славу и богатство - если бы я смог собрать его только с помощью NVRT C:

#include <stdio.h>

__global__ void do_stuff() { }

Я бы надеялся, что системные заголовки будут распознаны компилятором (времени выполнения), как и обычный компилятор, и что это будет «просто работать» (по модулю любой специфической для printf c машины). В качестве альтернативы, если бы это не сработало, я ожидал сообщения об ошибке об источнике stdio.h, недоступном с помощью вызова API «создания программы» (nvrtcCreateProgram()), поскольку я передача NULL и NULL в качестве последних двух аргументов.

Однако я получаю следующее:

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

Мне это кажется странным. Это означает, что компилятор времени выполнения может просматривать системные заголовки, но не может найти stddef.h, как nv cc или компилятор на стороне хоста.

Почему это происходит и каков идиоматический / рекомендуемый обходной путь?

Примечание: я бы хотел, чтобы обходной путь был кроссплатформенным, а не только работал на моем отдельном компьютере.

Answer 1

Дополнительный подход используется в библиотеке "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";

Если вы включаете эти строки в качестве заголовков с соответствующими именами в качестве ключей, вероятно, ваше ядро ​​будет компилироваться. использовать при компиляции ядра без использования NVRT C. Это тоже может быть полезно.

И последнее замечание: константы выше не определяют пространство выполнения __device__. Итак, либо вы добавляете туда __device__, либо указываете компилятору считать, что функции предназначены для выполнения только на устройстве, если не указано иное; это параметр компилятора --device-as-default-execution-space NVRT C.

Answer 2

Вот два решения, которые могут сработать, но я бы предпочел избегать . Если это, в конце концов, единственный разумный способ действия - прокомментируйте и скажите так:

1. Добавьте указанный c путь к stddef.h в качестве параметра компилятора (-I или --include-path=).
2. Передайте исходный код stddef.h вызову nvrtcCreateProgram().