[SYCL][RTC] Clarify and test handling of include paths

sycl-jit/jit-compiler/lib/rtc/DeviceCompilation.cpp

@@ -333,9 +333,10 @@ static void adjustArgs(const InputArgList &UserArgList,
   DAL.eraseArg(OPT_ftime_trace_granularity_EQ);
   DAL.eraseArg(OPT_ftime_trace_verbose);
 
-  for (auto *Arg : DAL) {
-    CommandLine.emplace_back(Arg->getAsString(DAL));
-  }
+  ArgStringList ASL;
+  for_each(DAL, [&DAL, &ASL](Arg *A) { A->render(DAL, ASL); });
+  transform(ASL, std::back_inserter(CommandLine),
+            [](const char *AS) { return std::string{AS}; });
 }
 
 static void setupTool(ClangTool &Tool, const std::string &DPCPPRoot,

sycl/doc/extensions/experimental/sycl_ext_oneapi_kernel_compiler.asciidoc

@@ -363,20 +363,11 @@ whose _Name_ is `foo/bar.h`.
 If such an entry is found, the compiler uses the associated _Content_ as the
 content of the include file.
 
-When the source language is `source_language::sycl`, the following header files
-are implicitly available.
-Therefore, the source string may `#include` these even without defining their
-content via the `include_files` property:
-
-* `<sycl/sycl.hpp>`;
-* The {cpp} standard library headers;
-* The SYCL backend headers `"sycl/backend/<backend_name>.hpp"` for any backends
-  that the implementation supports; and
-* Any SYCL extension headers in "sycl/ext" for extensions that the
-  implementation supports.
-
-The include files defined via the `include_files` property are searched first,
-before these implicitly available headers.
+[_Note_: This property is only required if an `#include` statement references a
+file that is not already implicitly available.
+For more information about implicitly available headers, see the section
+"Including files when the language is ``sycl``".
+_{endnote}_]
 
 _Effects (1):_ Creates a new `include_files` property with no (_Name_,
 _Content_) pairs.
@@ -655,6 +646,27 @@ _Throws:_
   `ext_oneapi_has_kernel(name)` returns `false`.
 |====
 
+=== Including files when the language is `sycl`
+
+When the source language is `source_language::sycl`, the compiler searches
+multiple locations to find files referenced by `#include` statements.
+Any include files defined via the `include_files` property are searched first,
+followed by the directories below, in order:
+
+1. The current working directory.
+2. Any directory added explicitly to the search list via the `build_options`
+   property.
+
+Finally, the compiler searches a set of implicitly available header files, which
+do not need to be specified via the `include_files` property:
+
+* `<sycl/sycl.hpp>`;
+* The {cpp} standard library headers;
+* The SYCL backend headers `"sycl/backend/<backend_name>.hpp"` for any backends
+  that the implementation supports; and
+* Any SYCL extension headers in `"sycl/ext"` for extensions that the
+  implementation supports.
+
 === Obtaining a kernel when the language is `sycl`
 
 When the kernel is defined in the language `source_language::sycl`, the host
@@ -938,3 +950,25 @@ There is no inherent reason why this functionality needs to be built into
 However, we don't yet have a utility library where this would go, and it may be
 hard for customers to discover this functionality if it is defined outside of
 this extension.
+
+== Non-normative implementation notes for {dpcpp}
+
+=== Supported `build_options` when the language is `sycl`
+
+The SYCL runtime compiler supports the following {dpcpp} options to be passed in
+the `build_options` property.
+
+[%header,cols="1,3"]
+|===
+|Option
+|Notes
+
+|`-I<dir>` +
+`-I <dir>` +
+`--include-directory=<dir>` +
+`--include-directory <dir>`
+| Add `<dir>` to to the search list for include files (see section "Including
+files when the language is ``sycl``"). This is useful, for example, to compile
+kernels using external libraries. Note that for the second and fourth form,
+`dir` is a separate element in the `build_options` list.
+|===

sycl/include/sycl/kernel_bundle.hpp

@@ -26,20 +26,22 @@
 #include <sycl/ext/oneapi/properties/property.hpp>       // build_options
 #include <sycl/ext/oneapi/properties/property_value.hpp> // and log
 
+#include <algorithm> // for copy
 #include <array>      // for array
 #include <cstddef>    // for std::byte
 #include <cstring>    // for size_t, memcpy
 #include <functional> // for function
-#include <iterator>   // for distance
+#include <iterator>   // for distance, back_inserter
 #include <memory>     // for shared_ptr, operator==, hash
 #if __has_include(<span>)
 #include <span>
 #endif
-#include <string>      // for string
-#include <type_traits> // for enable_if_t, remove_refer...
-#include <utility>     // for move
-#include <variant>     // for hash
-#include <vector>      // for vector
+#include <string>        // for string
+#include <type_traits>   // for enable_if_t, remove_refer...
+#include <unordered_map> // for unordered_map
+#include <utility>       // for move
+#include <variant>       // for hash
+#include <vector>        // for vector
 
 namespace sycl {
 inline namespace _V1 {
@@ -954,14 +956,19 @@ struct build_source_bundle_props;
 struct include_files
     : detail::run_time_property_key<include_files,
                                     detail::PropKind::IncludeFiles> {
-  include_files();
+  include_files() {}
   include_files(const std::string &name, const std::string &content) {
-    record.emplace_back(std::make_pair(name, content));
+    record.emplace(name, content);
   }
   void add(const std::string &name, const std::string &content) {
-    record.emplace_back(std::make_pair(name, content));
+    bool inserted = record.try_emplace(name, content).second;
+    if (!inserted) {
+      throw sycl::exception(make_error_code(errc::invalid),
+                            "Include file '" + name +
+                                "' is already registered");
+    }
   }
-  std::vector<std::pair<std::string, std::string>> record;
+  std::unordered_map<std::string, std::string> record;
 };
 using include_files_key = include_files;
 
@@ -977,8 +984,10 @@ struct build_options
     : detail::run_time_property_key<build_options,
                                     detail::PropKind::BuildOptions> {
   std::vector<std::string> opts;
+  build_options() {}
   build_options(const std::string &optsArg) : opts{optsArg} {}
   build_options(const std::vector<std::string> &optsArg) : opts(optsArg) {}
+  void add(const std::string &opt) { opts.push_back(opt); }
 };
 using build_options_key = build_options;
 
@@ -1116,7 +1125,10 @@ kernel_bundle<bundle_state::ext_oneapi_source> create_kernel_bundle_from_source(
     const std::string &Source, PropertyListT props = {}) {
   std::vector<std::pair<std::string, std::string>> IncludePairsVec;
   if constexpr (props.template has_property<include_files>()) {
-    IncludePairsVec = props.template get_property<include_files>().record;
+    const std::unordered_map<std::string, std::string> &IncludePairs =
+        props.template get_property<include_files>().record;
+    std::copy(IncludePairs.begin(), IncludePairs.end(),
+              std::back_inserter(IncludePairsVec));
   }
 
   return detail::make_kernel_bundle_from_source(SyclContext, Language, Source,
@@ -1132,7 +1144,10 @@ kernel_bundle<bundle_state::ext_oneapi_source> create_kernel_bundle_from_source(
     const std::vector<std::byte> &Bytes, PropertyListT props = {}) {
   std::vector<std::pair<std::string, std::string>> IncludePairsVec;
   if constexpr (props.template has_property<include_files>()) {
-    IncludePairsVec = props.template get_property<include_files>().record;
+    const std::unordered_map<std::string, std::string> &IncludePairs =
+        props.template get_property<include_files>().record;
+    std::copy(IncludePairs.begin(), IncludePairs.end(),
+              std::back_inserter(IncludePairsVec));
   }
 
   return detail::make_kernel_bundle_from_source(SyclContext, Language, Bytes,

sycl/test-e2e/KernelCompiler/include/A/header1.hpp

@@ -0,0 +1 @@
+#define DEFINE_1 fsA

sycl/test-e2e/KernelCompiler/include/B/header2.hpp

@@ -0,0 +1 @@
+#define DEFINE_2 fsB

sycl/test-e2e/KernelCompiler/include/C/header1.hpp

@@ -0,0 +1 @@
+#define DEFINE_1 fsC