Keyboard shortcuts

Press or to navigate between chapters

Press S or / to search in the book

Press ? to show this help

Press Esc to hide this help

CLI Flag Inventory

Quick Reference: 20 Most Important CUDA-Specific Flags

Flag (via -Xcudafe)nvcc EquivalentIDEffect
--diag_suppress=N--diag-suppress=N39Suppress diagnostic number N (comma-separated)
--diag_error=N--diag-error=N42Promote diagnostic N to error
--diag_warning=N--diag-warning=N41Demote diagnostic N to warning
--display_error_number--44Show #NNNNN-D error codes in output
--target=smXX--gpu-architecture=smXX245Set SM architecture target (parsed via sub_7525E0)
--relaxed_constexpr--expt-relaxed-constexpr104Allow constexpr cross-space calls
--extended-lambda--expt-extended-lambda106Enable __device__/__host__ __device__ lambdas in host code (dword_106BF38)
--device-c-rdc=true77Relocatable device code (separate compilation)
--keep-device-functions--keep-device-functions71Do not strip unused device functions
--no_warnings-w22Suppress all warnings
--promote_warnings-W23Promote all warnings to errors
--error_limit=N--32Maximum errors before abort (default: unbounded)
--force-lp64-m6465LP64 data model (pointer=8, long=8)
--output_mode=sarif--274SARIF JSON diagnostic output
--debug_mode-G82Full debug mode (sets 3 debug globals)
--device-syntax-only--72Device-side syntax check without codegen
--no-device-int128--52Disable __int128 on device
--zero_init_auto_vars--81Zero-initialize automatic variables
--fe-inlining--54Enable frontend inlining
--gen_c_file_name=path--45Set output .int.c file path

These are the flags most commonly passed through -Xcudafe for CUDA development. The full inventory of 276 flags follows below.


cudafe++ accepts 276 command-line flags registered in a flat table at dword_E80060. The flags are not parsed directly from the binary's argv -- NVIDIA's driver compiler nvcc decomposes its own options and invokes cudafe++ with the appropriate low-level flags. Users never run cudafe++ directly; instead, they pass options through nvcc -Xcudafe <flag>, which strips the -Xcudafe prefix and forwards the remainder as a bare argument to the cudafe++ process.

The flag system is implemented in three functions within cmd_line.c:

FunctionAddressLinesRole
register_command_flagsub_451F8025Insert one entry into the flag table
init_command_line_flagssub_4520103,849Register all 276 flags (called once)
proc_command_linesub_4596304,105Main parser: match argv against table, dispatch to 275-case switch
default_initsub_45EB40470Zero 350 global config variables + flag-was-set bitmap

Flag Table Structure

Each flag occupies a 40-byte entry in a contiguous array beginning at dword_E80060, with a maximum capacity of 552 entries (overflow triggers a panic via sub_40351D). The current count is tracked in dword_E80058.

struct flag_entry {                    // 40 bytes per entry
    int32_t   case_id;                 // dword_E80060[idx*10]    -- switch dispatch ID
    char*     name;                    // qword_E80068[idx*5]     -- long flag name string
    int16_t   short_char;              // word_E80070[idx*20]     -- single-char alias (0 if none)
    int8_t    is_valid;                // word_E80070[idx*20]+1   -- always 1
    int8_t    takes_value;             // byte_E80072[idx*40]     -- flag requires =<value> argument
    int32_t   visible;                 // dword_E80080[idx*10]    -- mode/action classification
    int8_t    is_boolean;              // byte_E80073[idx*40]     -- flag is on/off toggle
    int64_t   name_length;             // qword_E80078[idx*5]     -- strlen(name), precomputed
};

The flag-was-set bitmap at byte_E7FF40 spans 0x110 bytes (272 flag slots). When a flag is matched during parsing, the corresponding bit is set to record that the user explicitly provided it. This bitmap is zeroed by default_init before every compilation.

Registration Protocol

register_command_flag (sub_451F80) is called approximately 275 times from init_command_line_flags. Its prototype:

void register_command_flag(
    int    case_id,        // dispatch ID for the switch statement
    char*  name,           // "--name" (without the dashes)
    char   short_opt,      // single-letter alias, 0 for none
    char   takes_value,    // 1 if the flag requires =<value>
    int    mode_flag,      // visibility / classification
    char   enabled         // whether the flag is active
);

Some flags are registered as paired toggles -- --flag and --no_flag share the same case_id but set the target global to 1 or 0 respectively. These pairs are registered either by two calls to register_command_flag or by inline table population within init_command_line_flags.

Parsing Flow

proc_command_line (sub_459630) is the master CLI parser. It:

  1. Calls init_command_line_flags to populate the flag table (once)
  2. Allocates four hash tables for accumulating -D, -I, system include, and macro alias arguments
  3. Adjusts nine diagnostic severities by default via sub_4ED400: four are suppressed (severity 3: errors 1373, 1374, 1375, 2330) and five are demoted to remark (severity 4: errors 1257, 1633, 111, 185, 175)
  4. Enters the main loop over argv:
    • Scans for - prefix to identify flags
    • Handles -X short flags and --flag-name long flags
    • Handles --flag=value syntax via parse_flag_name_value (sub_451EC0)
    • Matches flag names against the registered table using strncmp against each entry's precomputed name_length
    • Dispatches to a giant switch(case_id) with 275 cases
  5. Executes post-parsing dialect resolution (described below)
  6. Opens output, error, and list files
  7. Treats the remaining non-flag argv entry as the input filename

The -Xcudafe Pass-Through

Users never invoke cudafe++ directly. The intended usage path is:

nvcc --some-option -Xcudafe --diag_suppress=1234 source.cu

nvcc strips -Xcudafe and passes --diag_suppress=1234 directly to the cudafe++ process as an argv element. Multiple -Xcudafe arguments accumulate. Because cudafe++ flags use -- long-form prefixes, there is no ambiguity with nvcc's own flag namespace.

Certain nvcc flags like --expt-extended-lambda and --expt-relaxed-constexpr are translated by nvcc into the corresponding cudafe++ internal flags (--extended-lambda, --relaxed_constexpr) before invocation. Users do not need to know the internal names.

Flag Catalog by Category

The 276 flags are grouped below by functional category. Each table lists:

  • ID -- the case_id used in the dispatch switch
  • Flag -- the --name as registered (paired flags shown as name / no_name)
  • Short -- single-character alias (dash required: -E, -C, etc.)
  • Arg -- whether the flag takes a =<value> argument
  • Effect -- what the flag does internally

Core EDG Flags (1--44)

These are standard Edison Design Group frontend options that predate NVIDIA's CUDA modifications.

IDFlagShortArgEffect
1strict-AnoEnable strict standards conformance mode
2strict_warnings-anoStrict mode with extra warnings
3no_line_commands-PnoSuppress #line directives in preprocessor output
4preprocess-EnoPreprocessor-only mode (output to stdout)
5comments-CnoPreserve comments in preprocessor output
6old_line_commands--noUse old-style # N "file" line directives
7old_c-KnoK&R C mode (calls set_c_mode(1))
8dependencies-MnoOutput #include dependency list (preprocessor-only)
9trace_includes-HnoPrint each #include file as it is opened
10il_display--noDump intermediate language after parsing
11anachronisms / no_anachronisms--noAllow/disallow anachronistic C++ constructs
12cfront_2.1-bnoCfront 2.1 compatibility mode
13cfront_3.0--noCfront 3.0 compatibility mode
14no_code_gen-nnoParse only, skip code generation
15signed_chars / unsigned_chars-snoDefault char signedness
16instantiate-tyesTemplate instantiation mode: none, all, used, local
17implicit_include / no_implicit_include-BnoEnable/disable implicit inclusion of template definitions
18suppress_vtbl / force_vtbl--noControl virtual table emission
19dollar-$noAllow $ in identifiers
20timing-#noPrint compilation phase timing
21version-vnoPrint version banner and continue
22no_warnings-wnoSuppress all warnings (sets severity threshold to error-only)
23promote_warnings-WnoPromote warnings to errors
24remarks-rnoEnable remark-level diagnostics
25c-mnoForce C language mode
26c++-pnoForce C++ language mode
27exceptions / no_exceptions-xnoEnable/disable C++ exception handling
28no_use_before_set_warnings-jnoSuppress "used before set" variable warnings
29include_directory-IyesAdd include search path (handles - for stdin)
30define_macro-DyesDefine preprocessor macro (builds linked list)
31undefine_macro-UyesUndefine preprocessor macro
32error_limit-eyesMaximum number of errors before abort
33list-LyesGenerate listing file
34xref-XyesGenerate cross-reference file
35error_output--yesRedirect error output to file
36output-oyesSet output file path
37db-dyesLoad debug database
38time_limit--yesSet compilation time limit
39diag_suppress--yesSuppress diagnostic numbers (comma-separated list)
40diag_remark--yesDemote diagnostics to remark severity
41diag_warning--yesSet diagnostics to warning severity
42diag_error--yesPromote diagnostics to error severity
43diag_once--yesEmit diagnostic only on first occurrence
44display_error_number / no_display_error_number--noShow/hide error code numbers in output

NVIDIA CUDA-Specific Flags (45--89)

These flags are NVIDIA additions absent from stock EDG. They control CUDA compilation modes, device code generation, and host/device interaction.

IDFlagArgEffect
45gen_c_file_nameyesSet output .int.c file path (qword_106BF20)
46msvc_target_versionyesMSVC version for compatibility (dword_126E1D4)
47host-stub-linkage-explicitnoUse explicit linkage on host stubs
48static-host-stubnoGenerate static host stubs
49device-hidden-visibilitynoApply hidden visibility to device symbols
50no-hidden-visibility-on-unnamed-nsnoExempt unnamed namespaces from hidden visibility
51no-multiline-debugnoDisable multiline debug info
52no-device-int128noDisable __int128 on device
53no-device-float128noDisable __float128 on device
54fe-inliningnoEnable frontend inlining (dword_106C068 = 1)
55modify-stack-limityesControl stack limit modification (dword_106C064)
56fassociative-mathnoEnable associative floating-point math
57orig_src_file_nameyesOriginal source file name (before preprocessing)
58orig_src_path_nameyesOriginal source path name (full path)
59frandom-seedyesRandom seed for reproducible output
60check-template-param-qualnoCheck template parameter qualifications
61check-clock-callnoValidate clock() calls in device code
62check-ffs-callnoValidate ffs() calls in device code
63check-routine-address-takennoCheck when device routine address is taken
64check-memory-clobbernoValidate memory clobber in inline asm
65force-lp64noLP64 data model: pointer=8, long=8
66force-llp64noLLP64 data model: pointer=4, long=4
67pgi_llvmnoPGI/LLVM backend mode
68pgi_arch_ppcnoPGI PowerPC architecture
69pgi_arch_aarch64noPGI AArch64 architecture
70pgi_versionyesPGI compiler version number
71keep-device-functionsnoDo not strip unused device functions
72device-syntax-onlynoDevice-side syntax check without codegen
73device-time-tracenoEnable device compilation time tracing
74force_linkonce_to_weaknoConvert linkonce to weak linkage
75disable_host_implicit_call_checknoSkip implicit call validation on host
76no_strict_cuda_errornoRelax strict CUDA error checking
77device-cnoRelocatable device code (RDC) mode
78no-shadow-functionsnoDisable function shadowing in device code
79disable_ext_lambda_cachenoDisable extended lambda capture cache
80no-constant-variable-inferencingnoDisable constexpr variable inference on device
81zero_init_auto_varsnoZero-initialize automatic variables
82debug_modenoFull debug mode (sets 3 debug globals to 1)
83gen_module_id_filenoGenerate module ID file
84include_file_nameyesForced include file name
85gen_device_file_nameyesDevice-side output file name
86stub_file_nameyesStub file output path
87module_id_file_nameyesModule ID file path
88tile_bc_file_nameyesTile bitcode file path
89tile-onlynoTile-only compilation mode

Architecture and Host Compiler Flags (90--114)

These flags identify the target architecture and host compiler for compatibility emulation.

IDFlagShortArgEffect
90m32--no32-bit mode: pointer=4, long=4, all types sized for ILP32
91m64--no64-bit mode (default on Linux x86-64)
92Version-VnoPrint version with different copyright format, then exit(1)
93compiler_bindir--yesHost compiler binary directory
94sdk_dir--yesSDK directory path
95pgc++--noPGI C++ compiler mode
96icc--noIntel ICC compiler mode
97icc_version--yesIntel ICC version number
98icx--noIntel ICX (oneAPI) compiler mode
99grco--noGRCO compiler mode
100allow_managed--noAllow __managed__ variable declarations
101gen_system_templates_from_text--noGenerate system templates from text
102no_host_device_initializer_list--noDisable HD initializer_list support
103no_host_device_move_forward--noDisable HD std::move/std::forward
104relaxed_constexpr--noRelaxed constexpr rules for device code (--expt-relaxed-constexpr)
105dont_suppress_host_wrappers--noEmit host wrapper functions unconditionally
106arm_cross_compiler--noARM cross-compilation mode
107target_woa--noWindows on ARM target
108gen_div_approx_no_ftz--noGenerate approximate division without flush-to-zero
109gen_div_approx_ftz--noGenerate approximate division with flush-to-zero
110shared_address_immutable--noShared memory addresses are immutable
111uumn--noUnnamed union member naming

C++ Language Feature Toggle Flags (115--275)

The largest group -- approximately 120 paired boolean toggles that control individual C++ language features. Most are inherited from EDG's configuration surface. Each pair shares a case_id and sets a global variable to 1 (--flag) or 0 (--no_flag).

Precompiled Headers (115--121)

IDFlagArgEffect
115unsigned_wchar_tnowchar_t is unsigned
116create_pchyesCreate precompiled header file
117use_pchyesUse existing precompiled header
118pchnoEnable PCH mode
119pch_messages / no_pch_messagesnoShow/hide PCH status messages
120pch_verbose / no_pch_verbosenoVerbose PCH output
121pch_diryesPCH file directory

Core C++ Feature Toggles (122--170)

IDFlagArgDefault
122restrict / no_restrictnoon
123long_lifetime_temps / short_lifetime_tempsno--
124wchar_t_keyword / no_wchar_t_keywordnoon
125pack_alignmentyes--
126alternative_tokens / no_alternative_tokensnoon
127svr4 / no_svr4no--
128brief_diagnostics / no_brief_diagnosticsno--
129nonconst_ref_anachronism / no_nonconst_ref_anachronismno--
130no_preproc_onlyno--
131rtti / no_rttinoon
132building_runtimeno--
133bool / no_boolnoon
134array_new_and_delete / no_array_new_and_deleteno--
135explicit / no_explicitno--
136namespaces / no_namespacesnoon
137using_std / no_using_stdno--
138remove_unneeded_entities / no_remove_unneeded_entitiesnoon
139typename / no_typenameno--
140implicit_typename / no_implicit_typenamenoon
141special_subscript_cost / no_special_subscript_costno--
143old_style_preprocessingno--
144old_for_init / new_for_initno--
145for_init_diff_warning / no_for_init_diff_warningno--
146distinct_template_signatures / no_distinct_template_signaturesno--
147guiding_decls / no_guiding_declsnoon
148old_specializations / no_old_specializationsnoon
149wrap_diagnostics / no_wrap_diagnosticsno--
150implicit_extern_c_type_conversion / no_implicit_extern_c_type_conversionno--
151long_preserving_rules / no_long_preserving_rulesno--
152extern_inline / no_extern_inlineno--
153multibyte_chars / no_multibyte_charsno--
154embedded_c++noEmbedded C++ mode
155vla / no_vlano--
156enum_overloading / no_enum_overloadingno--
157nonstd_qualifier_deduction / no_nonstd_qualifier_deductionno--
158late_tiebreaker / early_tiebreakerno--
159preincludeyes--
160preinclude_macrosyes--
161pending_instantiationsyes--
162const_string_literals / no_const_string_literalsnoon
163class_name_injection / no_class_name_injectionnoon
164arg_dep_lookup / no_arg_dep_lookupnoon
165friend_injection / no_friend_injectionnoon
166nonstd_using_decl / no_nonstd_using_declno--
168designators / no_designatorsno--
169extended_designators / no_extended_designatorsno--
170variadic_macros / no_variadic_macrosno--
171extended_variadic_macros / no_extended_variadic_macrosno--

Include Paths and Module Support (167, 172, 256--265)

Note: These flags use non-contiguous IDs because sys_include and incl_suffixes are registered early, while the C++20 module flags use a separate ID range (256+).

IDFlagArgEffect
167sys_includeyesSystem include directory
172incl_suffixesyesInclude file suffix list (default "::stdh:")
256modules_directoryyesC++20 modules directory
257ms_mod_file_mapyesMSVC module file mapping
258ms_header_unityesMSVC header unit
259ms_header_unit_quoteyesMSVC quoted header unit
260ms_header_unit_angleyesMSVC angle-bracket header unit
261ms_mod_interface / no_ms_mod_interfacenoMSVC module interface mode
262ms_internal_partition / no_ms_internal_partitionnoMSVC internal partition mode
263ms_translate_include / no_ms_translate_includenoMSVC translate #include to import
264modules / no_modulesnoEnable/disable C++20 modules
265module_import_diagnostics / no_module_import_diagnosticsnoModule import diagnostic messages

Host Compiler and Language Feature Toggles (182--239)

Note: All IDs below are verified against the decompiled init_command_line_flags (sub_452010). Flags are registered by sub_451F80 (explicit call) or by inline array population. IDs are not sequential -- gaps exist where flags were removed or repurposed.

IDFlagArgDefault
182gcc / no_gccnoGCC compatibility mode
183g++ / no_g++noG++ mode (alias for GCC C++ mode)
184gnu_versionyesGCC version number (default 80100 = GCC 8.1.0)
185report_gnu_extensionsnoReport use of GNU extensions
186short_enums / no_short_enumsnoUse minimal-size enum representation
187clang / no_clangnoClang compatibility mode
188clang_versionyesClang version number (default 90100 = Clang 9.1.0)
189strict_gnu / no_strict_gnunoStrict GNU mode
190db_nameyesDebug database name
191long_longnoAllow long long type
192context_limityesMaximum template instantiation context depth
193set_flag / clear_flagyesRaw flag manipulation via off_D47CE0 lookup table
194edg_base_diryesEDG base directory (error on invalid path)
195embedded_c / no_embedded_cnoEmbedded C mode (not relevant to CUDA)
196thread_local_storage / no_thread_local_storagenothread_local support
197trigraphs / no_trigraphsnoTrigraph processing (default on)
198nonstd_default_arg_deduction / no_nonstd_default_arg_deductionno--
199stdc_zero_in_system_headers / no_stdc_zero_in_system_headersno--
200template_typedefs_in_diagnostics / no_template_typedefs_in_diagnosticsno--
202uliterals / no_uliteralsnoUnicode literals (u"", U"", u8"")
203type_traits_helpers / no_type_traits_helpersnoIntrinsic type traits
204c++11 / c++0xnoC++11 mode (sets dword_126EF68 to 201103 or 199711)
205list_macrosnoList all defined macros after preprocessing
206dump_configurationnoDump full compiler configuration
207dump_legacy_as_targetyesDump legacy configuration in target format
208signed_bit_fields / unsigned_bit_fieldsnoDefault bit-field signedness
210check_concatenations / no_check_concatenationsnoString literal concatenation checks
211unicode_source_kindyesSource encoding: UTF-8=1, UTF-16LE=2, UTF-16BE=3, none=0
212lambdas / no_lambdasnoC++ lambda expressions
213rvalue_refs / no_rvalue_refsnoRvalue references
214rvalue_ctor_is_copy_ctor / rvalue_ctor_is_not_copy_ctornoRvalue constructor treatment
215gen_move_operations / no_gen_move_operationsnoImplicit move constructor/assignment (default on)
216auto_type / no_auto_typenoC++11 auto type deduction
217auto_storage / no_auto_storagenoauto as storage class (C++03 meaning)
218nonstd_instantiation_lookup / no_nonstd_instantiation_lookupno--
219nullptr / no_nullptrnonullptr keyword
220gcc89_inliningnoGCC 8.9-era inlining behavior
221nonstd_gnu_keywords / no_nonstd_gnu_keywordsnoGNU extension keywords
222default_nocommon_tentative_definitions / default_common_tentative_definitionsnoTentative definition linkage
223no_token_separators_in_pp_outputno--
224c23_typeof / no_c23_typeofnoC23 typeof operator
225c++11_sfinae / no_c++11_sfinaenoC++11 SFINAE rules
226c++11_sfinae_ignore_access / no_c++11_sfinae_ignore_accessnoIgnore access checks in SFINAE
227variadic_templates / no_variadic_templatesnoParameter packs and pack expansion
228c++03noC++03 mode (sets dword_126EF68 to 199711)
229func_prototype_tags / no_func_prototype_tagsno--
230implicit_noexcept / no_implicit_noexceptnoImplicit noexcept on destructors
231unrestricted_unions / no_unrestricted_unionsnoUnrestricted unions (C++11)
232max_depth_constexpr_callyesMaximum constexpr recursion depth (default 200)
233max_cost_constexpr_callyesMaximum constexpr evaluation cost (default 256)
234delegating_constructors / no_delegating_constructorsno--
235lossy_conversion_warning / no_lossy_conversion_warningno--
236deprecated_string_conv / no_deprecated_string_convnoDeprecated string literal to char* conversion
237user_defined_literals / no_user_defined_literalsnoUDL support
238preserve_lvalues_with_same_type_casts / no_...no--
239nonstd_anonymous_unions / no_nonstd_anonymous_unionsno--

Late C++/Architecture/Output Flags (240--258)

IDFlagArgEffect
240c++14noC++14 mode (sets dword_126EF68 to 201402)
241c11noC11 mode (sets dword_126EF68 to 201112)
242c17noC17 mode (sets dword_126EF68 to 201710)
243c23noC23 mode (sets dword_126EF68 to 202311)
244digit_separators / no_digit_separatorsnoC++14 digit separators (1'000'000)
245targetyesSM architecture string, parsed via sub_7525E0 into dword_126E4A8
246c++17noC++17 mode (sets dword_126EF68 to 201703)
247utf8_char_literals / no_utf8_char_literalsnoUTF-8 character literal support
248stricter_template_checkingnoAdditional template constraint checks
249exc_spec_in_func_type / no_exc_spec_in_func_typenoException spec as part of function type (C++17)
250aligned_new / no_aligned_newnoAligned operator new (C++17)
251c++20noC++20 mode (sets dword_126EF68 to 202002)
252c++23noC++23 mode (sets dword_126EF68 to 202302)
253ms_std_preprocessor / no_ms_std_preprocessornoMSVC standard preprocessor mode
268partial-linknoPartial linking mode
273dump_command_optionsnoPrint all registered flag names
274output_modeyesOutput format: text (0) or sarif (1)
275incognito / no_incognitonoIncognito mode

Note: Many IDs in the 240-252 range serve double duty as both C/C++ standard selectors and feature toggles. The standard selection IDs are also cross-referenced in the Language Standard Selection section above.

Inline-Registered Paired Flags

Seven additional paired flags are registered through inline table population rather than calls to register_command_flag. They share the same entry structure but are populated directly into the array:

FlagEffect
relaxed_abstract_checking / no_relaxed_abstract_checkingRelax abstract class checks
concepts / no_conceptsC++20 concepts support
colors / no_colorsColorized diagnostic output
keep_restrict_in_signatures / no_keep_restrict_in_signaturesPreserve restrict in mangled names
check_unicode_security / no_check_unicode_securityUnicode security checks (homoglyph detection)
old_id_chars / no_old_id_charsLegacy identifier character rules
add_match_notes / no_add_match_notesAdd notes about matching overloads

Language Standard Selection

Six language standard flags set dword_126EF68 (the internal __cplusplus / __STDC_VERSION__ value) and trigger corresponding mode changes:

C Standards

IDFlagdword_126EF68 valueC standard
7old_c(K&R)Pre-ANSI C via set_c_mode(1)
179c89198912ANSI C / C89
178c99199901C99
241c11201112C11
242c17201710C17
243c23202311C23

C++ Standards

IDFlagdword_126EF68 valueC++ standard
228c++03199711C++98/03 (also aliased as c++98 via --c++11 flag ID 204 with conditional)
204c++11201103C++11 (sets 199711 if dword_E7FF14 is unset or C mode)
240c++14201402C++14
246c++17201703C++17
251c++20202002C++20
252c++23202302C++23

When a C++ standard is selected, the post-parsing dialect resolution logic automatically enables the corresponding feature flags. For example, selecting --c++11 (value 201103) enables lambdas, rvalue references, auto type deduction, nullptr, variadic templates, and other C++11 features. The resolution logic also interacts with GCC/Clang version thresholds to determine which extensions are available.

Diagnostic Control Flags

The five diag_* flags (IDs 39--43) accept comma-separated lists of diagnostic numbers. The parser strips whitespace, splits on commas, and calls sub_4ED400(number, severity, 1) for each number:

--diag_suppress=1234,5678       # suppress errors 1234 and 5678
--diag_warning=20001            # demote CUDA error 20001 to warning
--diag_error=111                # promote diagnostic 111 to error
--diag_remark=185               # demote diagnostic 185 to remark
--diag_once=175                 # emit diagnostic 175 only once

The error number system is documented in Diagnostic System Overview. Numbers above 3456 in the internal range correspond to the 20000-series CUDA errors via the offset formula display_code = internal_code + 16543.

Post-Parsing Dialect Resolution

After the main parsing loop completes, proc_command_line executes a large block of dialect resolution logic that:

  1. Resolves host compiler mode conflicts -- If both --gcc and --clang are set, or --cfront_2.1 is combined with modern modes, the resolution picks one and adjusts feature flags accordingly
  2. Sets C++ feature flags from __cplusplus version -- Based on the value in dword_126EF68:
    • 199711 (C++98/03): baseline features only
    • 201103 (C++11): enables lambdas, rvalue refs, auto, nullptr, variadic templates, range-based for, delegating constructors, unrestricted unions, user-defined literals
    • 201402 (C++14): adds digit separators, generic lambdas, relaxed constexpr
    • 201703 (C++17): adds aligned new, exception spec in function type, structured bindings
    • 202002 (C++20): adds concepts, modules, coroutines
    • 202302 (C++23): adds latest features
  3. Applies GCC version thresholds -- When in GCC compatibility mode, certain features are gated on the GCC version number stored in qword_126EF98 (default 80100 = GCC 8.1.0). Known thresholds:
    • 40299 (0x9D6B): GCC 4.2
    • 40599 (0x9E97): GCC 4.5
    • 40699 (0x9EFB): GCC 4.6
    • Higher versions enable progressively more features
  4. Opens output files -- Error output, listing file, output file
  5. Processes the input filename -- The remaining non-flag argv entry

Key Globals After Resolution

GlobalTypeContent
dword_126EF68int32__cplusplus / __STDC_VERSION__ value
dword_126EFB4int32Language mode: 0=unset, 1=C, 2=C++
dword_126EFA8int32GCC compatibility enabled
dword_126EFA4int32Clang compatibility enabled
qword_126EF98int64GCC version (default 80100)
qword_126EF90int64Clang version (default 90100)
dword_126EFB0int32GNU extensions enabled
dword_126EFACint32Clang extensions enabled
dword_126E4A8int32SM architecture code (from --target)
dword_126E1D4int32MSVC target version

The set_flag / clear_flag Mechanism

Flag ID 199 (--set_flag / --clear_flag) provides a raw escape hatch. The argument is a flag name looked up in the off_D47CE0 table -- an array of {name, global_address} pairs. If the name is found, the corresponding global variable is set to the provided integer value (--set_flag=name=value) or cleared to 0 (--clear_flag=name). This mechanism allows nvcc to toggle internal EDG configuration flags that do not have dedicated CLI flag registrations.

Default Values

default_init (sub_45EB40) runs before proc_command_line and initializes approximately 350 global configuration variables. Notable non-zero defaults:

GlobalDefaultMeaning
dword_106C2101Exceptions enabled
dword_106C1801RTTI enabled
dword_106C1781bool is keyword
dword_106C1941Namespaces enabled
dword_106C19C1Argument-dependent lookup enabled
dword_106C1A01Class name injection enabled
dword_106C1A41String literals are const
dword_106C1881wchar_t is keyword
dword_106C18C1Alternative tokens enabled
dword_106C1401Compound literals allowed
dword_106C1381Dependent name processing enabled
dword_106C1341Template parsing enabled
dword_106C12C1Friend injection enabled
dword_106BDB81restrict enabled
dword_106BDB01Remove unneeded entities enabled
dword_106BD981Trigraphs enabled
dword_106BD681Guiding declarations allowed
dword_106BD581Old specializations allowed
dword_106BD541Implicit typename enabled
dword_106BE841Generate move operations enabled
dword_106C0641Stack limit modification enabled
qword_106BD10200Max constexpr recursion depth
qword_106BD08256Max constexpr evaluation cost
qword_126EF9880100Default GCC version (8.1.0)
qword_126EF9090100Default Clang version (9.1.0)
qword_126EF781926MSVC version threshold
qword_126EF7099999Some upper bound sentinel

Conflict Detection

Before the main parsing loop, check_conflicting_flags (sub_451E80) verifies that flags 3, 193, 194, and 195 (no_line_commands, set_flag, clear_flag, and related flags) are not used in conflicting combinations. If any conflict is detected, error 1027 is emitted.

Version Banners

Two flags print version information:

--version (ID 21, -v):

cudafe: NVIDIA (R) Cuda Language Front End
Portions Copyright (c) 2005, 2024 NVIDIA Corporation
Portions Copyright (c) 1988-2018, 2024 Edison Design Group Inc.
Based on Edison Design Group C/C++ Front End, version 6.6
Cuda compilation tools, release 13.0, V13.0.88

--Version (ID 92, -V): Prints a different copyright format with full date/time stamp, then calls exit(1).

Cross-References