[SYCL][Doc] Move internal design docs (#5556)

Move all the internal design documents to a separate directory.  This
declutters the "sycl/doc" directory, making it easier to find the
end-user documentation.  It also moves several documents out of
"sycl/doc/extensions" which were not actually extension specifications.

We decided that the "sycl/doc/dev" directory should hold BKMs (best
known methods) for developers, so rename it accordingly.

Author: gmlueck

sycl/ReleaseNotes.md

@@ -15,18 +15,18 @@ Release notes for commit range 4fc5ebe..bd68232
  - Added support for SYCL 2020 exceptions [5c0f748][eef07606][5af8c43d]
  - Added support for [SYCL_EXT_INTEL_BF16_CONVERSION extension](doc/extensions/experimental/SYCL_EXT_INTEL_BF16_CONVERSION.asciidoc)
    [8075463]
- - Added support for fallback implementation of [assert feature](doc/Assert.md)
+ - Added support for fallback implementation of [assert feature](doc/design/Assert.md)
    [56c9ec4]
  - Added support SYCL 2020 `sycl::logical_and` and `sycl::logical_or` operators
    [6c077a0]
 ### Documentation
- - Added design document for [optional kernel features](doc/OptionalDeviceFeatures.md)
+ - Added design document for [optional kernel features](doc/design/OptionalDeviceFeatures.md)
    [88cfe16]
  - Added [SYCL_INTEL_bf16_conversion extension document](doc/extensions/experimental/SYCL_EXT_INTEL_BF16_CONVERSION.asciidoc)
    [9f8cc3af]
  - Align [SYCL_EXT_ONEAPI_GROUP_MASK extension](doc/extensions/supported/SYCL_EXT_ONEAPI_SUB_GROUP_MASK.asciidoc)
    with SYCL 2020 specification [a06bd1fb]
- - Added [documentation](doc/SYCLInstrumentationUsingXPTI.md) of XPTI related
+ - Added [documentation](doc/design/SYCLInstrumentationUsingXPTI.md) of XPTI related
    tracing in SYCL [1308fe7b]
  - Align `SYCL_EXT_ONEAPI_LOCAL_MEMORY` extension
    [document](doc/extensions/supported/SYCL_EXT_ONEAPI_LOCAL_MEMORY.asciidoc) with SYCL 2020
@@ -1104,7 +1104,7 @@ Release notes for commit range c9d50752..5d7e0925
   - The SYCL library doesn't guarantee stable API/ABI, so applications compiled
     with older version of the SYCL library may not work with new one.
     The workaround is to rebuild the application.
-    [ABI policy guide](doc/ABIPolicyGuide.md)
+    [ABI policy guide](doc/developer/ABIPolicyGuide.md)
   - Using `cl::sycl::program` API to refer to a kernel defined in another
     translation unit leads to undefined behavior
   - Linkage errors with the following message:
@@ -1243,7 +1243,7 @@ Release notes for commit range 5976ff0..1fc0e4f
   - The SYCL library doesn't guarantee stable API/ABI, so applications compiled
     with older version of the SYCL library may not work with new one.
     The workaround is to rebuild the application.
-    [ABI policy guide](doc/ABIPolicyGuide.md)
+    [ABI policy guide](doc/developer/ABIPolicyGuide.md)
   - Using `cl::sycl::program` API to refer to a kernel defined in another
     translation unit leads to undefined behavior
   - Linkage errors with the following message:
@@ -1431,7 +1431,7 @@ Release notes for the commit range 75b3dc2..5976ff0
   - The SYCL library doesn't guarantee stable API/ABI, so applications compiled
     with older version of the SYCL library may not work with new one.
     The workaround is to rebuild the application.
-    [ABI policy guide](doc/ABIPolicyGuide.md)
+    [ABI policy guide](doc/developer/ABIPolicyGuide.md)
   - Using `cl::sycl::program` API to refer to a kernel defined in another
     translation unit leads to undefined behavior
   - Linkage errors with the following message:
@@ -1585,7 +1585,7 @@ Release notes for the commit range ba404be..24726df
   - The SYCL library doesn't guarantee stable API/ABI, so applications compiled
     with older version of the SYCL library may not work with new one.
     The workaround is to rebuild the application.
-    [ABI policy guide](doc/ABIPolicyGuide.md)
+    [ABI policy guide](doc/developer/ABIPolicyGuide.md)
   - Using `cl::sycl::program` API to refer to a kernel defined in another
     translation unit leads to undefined behavior
   - Linkage errors with the following message:
@@ -1661,7 +1661,7 @@ Release notes for the commit range ba404be..67d3d9e
   - Updated prerequisites in GetStartedGuide(doc/GetStartedGuide.md) [5d0d034]
   - Published a [proposal](doc/extensions/KernelRHSAttributes/SYCL_INTEL_attribute_style.asciidoc)
     for function-type attributes (right-sided) for kernel attributes [5d5351b]
-  - The [compiler and runtime design doc](doc/CompilerAndRuntimeDesign.md) has
+  - The [compiler and runtime design doc](doc/design/CompilerAndRuntimeDesign.md) has
     been updated to describe the CUDA target and reflect changed action graphs
     [91b597b] [212a26c]
   - [ExtendedAtomics documentation](doc/extensions/ExtendedAtomics/README.md)
@@ -1673,7 +1673,7 @@ Release notes for the commit range ba404be..67d3d9e
   - Published [device_specific_kernel_queries](doc/extensions/DeviceSpecificKernelQueries/SYCL_INTEL_device_specific_kernel_queries.asciidoc)
     extension which rephrases work group queries as device-specific kernel
     queries [4c07ff8]
-  - Added more information about the [plugin interface (PI)](doc/PluginInterface.md)
+  - Added more information about the [plugin interface (PI)](doc/design/PluginInterface.md)
     [0614e9a]
   - [Contribution guidelines](../CONTRIBUTING.md) were simplified, now sign-off
     line is not required [7886fd8]
@@ -1736,7 +1736,7 @@ Release notes for the commit range ba404be..67d3d9e
   - The SYCL library doesn't guarantee stable API/ABI, so applications compiled
     with older version of the SYCL library may not work with new one.
     The workaround is to rebuild the application.
-    [ABI policy guide](doc/ABIPolicyGuide.md)
+    [ABI policy guide](doc/developer/ABIPolicyGuide.md)
   - Using `cl::sycl::program` API to refer to a kernel defined in another
     translation unit leads to undefined behavior
   - Linkage errors with the following message:

sycl/doc/FAQ.md

@@ -58,7 +58,7 @@ compiler command-line interface in general, please refer to the
 To learn more about the concepts behind this flow, and the DPC++
 Compiler internals as such, we welcome you to study our
 [DPC++ Compiler and Runtime architecture design](
-CompilerAndRuntimeDesign.md) document.
+design/CompilerAndRuntimeDesign.md) document.
 
 
 ## Using applications built with DPC++
@@ -150,7 +150,7 @@ debugging tools of your choice for the host device code.
 
 Furthermore, developers can extend capabilities of the DPC++ Runtime to
 non-OpenCL devices by writing correspondent plugins. To learn more, please
-check out our [Plugin Interface Guide](PluginInterface.md).
+check out our [Plugin Interface Guide](design/PluginInterface.md).
 
 ### Q: DPC++ applications hang on Intel GPUs while working well on other devices
 **A:** One of the common reasons is Intel GPUs feature called "hang check".

sycl/doc/GetStartedGuide.md

@@ -54,7 +54,7 @@ docker run --name sycl_build -it -v /local/workspace/dir/:/src ghcr.io/intel/llv
 ```
 
 This command will start a terminal session, from which you can proceed with the
-instructions below. See [Docker BKMs](dev/DockerBKMs.md) for more info on Docker
+instructions below. See [Docker BKMs](developer/DockerBKMs.md) for more info on Docker
 commands.
 
 ### Create DPC++ workspace
@@ -434,7 +434,7 @@ command:
 
 ### Obtain prerequisites for ahead of time (AOT) compilation
 
-[Ahead of time compilation](CompilerAndRuntimeDesign.md#ahead-of-time-aot-compilation)
+[Ahead of time compilation](design/CompilerAndRuntimeDesign.md#ahead-of-time-aot-compilation)
 requires ahead of time compiler available in `PATH`. There is
 AOT compiler for each device type:
 
@@ -660,7 +660,7 @@ clang++ -fsycl -fsycl-targets=spir64_gen,spir64_x86_64 simple-sycl-app.cpp -o si
 
 Additionally, user can pass specific options of AOT compiler to
 the DPC++ compiler using ```-Xsycl-target-backend``` option, see
-[Device code formats](CompilerAndRuntimeDesign.md#device-code-formats) for
+[Device code formats](design/CompilerAndRuntimeDesign.md#device-code-formats) for
 more. To find available options, execute:
 
 ```ocloc compile --help``` for GPU,

sycl/doc/conf.py

@@ -46,13 +46,33 @@
 # The suffix of source filenames.
 source_suffix = ['.rst', '.md']
 
-# Extensions are mostly in asciidoc which has poor support in Sphinx
-exclude_patterns = ['extensions/*']
+exclude_patterns = [
+    # Extensions are mostly in asciidoc which has poor support in Sphinx.
+    'extensions/*',
+
+    # Sphinx complains about syntax errors in these files.
+    'design/DeviceLibExtensions.rst',
+    'design/SYCLPipesLoweringToSPIRV.rst',
+    'design/fpga_io_pipes_design.rst',
+    'design/Reduction_status.md'
+]
 
 suppress_warnings = [ 'misc.highlighting_failure' ]
 
 def on_missing_reference(app, env, node, contnode):
-    new_target = "https://github.com/intel/llvm/tree/sycl/sycl/doc/" + node['reftarget']
+    # Get the directory that contains the *source* file of the link.  These
+    # files are always relative to the directory containing "conf.py"
+    # (<top>/sycl/doc).  For example, the file "sycl/doc/design/foo.md" will
+    # have a directory "design".
+    refdoc_components = node['refdoc'].split('/')
+    dirs = '/'.join(refdoc_components[:-1])
+    if dirs: dirs += '/'
+
+    # A missing reference usually occurs when the target file of the link is
+    # not processed by Sphinx.  Compensate by creating a link that goes to the
+    # file's location in the GitHub repo.
+    new_target = "https://github.com/intel/llvm/tree/sycl/sycl/doc/" + dirs + \
+        node['reftarget']
 
     newnode = nodes.reference('', '', internal=False, refuri=new_target)
     newnode.append(contnode)

sycl/doc/Assert.md renamed to sycl/doc/design/Assert.md

@@ -41,7 +41,7 @@ int main() {
 In this use-case every work-item with even index along 0 dimension will trigger
 assertion failure. Assertion failure should trigger a call to `std::abort()` at
 host as described in
-[extension](extensions/Assert/SYCL_INTEL_ASSERT.asciidoc).
+[extension](../extensions/supported/SYCL_EXT_ONEAPI_ASSERT.asciidoc).
 Even though multiple failures of the same or different assertions can happen in
 multiple work-items, implementation is required to deliver at least one
 assertion. The assertion failure message is printed to `stderr` by DPCPP
@@ -81,7 +81,7 @@ practical cases.
 ## How it works?
 
 `assert(expr)` macro ends up in call to `__devicelib_assert_fail`. This function
-is part of [Device library extension](extensions/C-CXX-StandardLibrary/DeviceLibExtensions.rst#cl_intel_devicelib_cassert).
+is part of [Device library extension](DeviceLibExtensions.rst#cl_intel_devicelib_cassert).
 
 The format of the assert message is unspecified, but it will always include the
 text of the failing expression, the values of the standard macros `__FILE__` and

sycl/doc/CompileTimeProperties.md renamed to sycl/doc/design/CompileTimeProperties.md

@@ -6,7 +6,7 @@ specifying properties which are known at compile time.  This extension is not
 itself a feature, but rather a building block that can be incorporated into
 other features.
 
-[1]: <extensions/proposed/SYCL_EXT_ONEAPI_PROPERTIES.asciidoc>
+[1]: <../extensions/proposed/SYCL_EXT_ONEAPI_PROPERTIES.asciidoc>
 
 There are a number of situations where we plan to use compile-time constant
 properties, but this design document does not attempt to address them all.
@@ -40,7 +40,7 @@ One use for compile-time properties is with types that are used exclusively
 for declaring global variables.  One such example is the
 [SYCL\_EXT\_ONEAPI\_DEVICE\_GLOBAL][2] extension:
 
-[2]: <extensions/proposed/SYCL_EXT_ONEAPI_DEVICE_GLOBAL.asciidoc>
+[2]: <../extensions/proposed/SYCL_EXT_ONEAPI_DEVICE_GLOBAL.asciidoc>
 
 ```
 namespace sycl::ext::oneapi {
@@ -148,7 +148,7 @@ kernel arguments.  For example, the [SYCL\_ONEAPI\_accessor\_properties][6]
 extension could be redesigned to use compile-time properties.  Such a redesign
 might look like:
 
-[6]: <extensions/supported/SYCL_EXT_ONEAPI_ACCESSOR_PROPERTIES.asciidoc>
+[6]: <../extensions/supported/SYCL_EXT_ONEAPI_ACCESSOR_PROPERTIES.asciidoc>
 
 ```
 namespace sycl {
@@ -274,7 +274,7 @@ the [SYCL\_EXT\_ONEAPI\_PROPERTIES][8] extension.  There are two ways the
 application can specify these properties.  The first is by passing a
 `property_list` parameter to the function that submits the kernel:
 
-[8]: <extensions/proposed/SYCL_EXT_ONEAPI_PROPERTIES.asciidoc>
+[8]: <../extensions/proposed/SYCL_EXT_ONEAPI_PROPERTIES.asciidoc>
 
 ```
 namespace sycl {

sycl/doc/CompilerAndRuntimeDesign.md renamed to sycl/doc/design/CompilerAndRuntimeDesign.md

@@ -951,4 +951,4 @@ with any other address space (including default).
 
 ## DPC++ Language extensions to SYCL
 
-List of language extensions can be found at [extensions](extensions)
+List of language extensions can be found at [extensions](../extensions)

sycl/doc/DeviceGlobal.md renamed to sycl/doc/design/DeviceGlobal.md

@@ -4,7 +4,7 @@ This document describes the implementation design for the DPC++ extension
 [SYCL\_EXT\_ONEAPI\_DEVICE\_GLOBAL][1], which allows applications to declare
 global variables in device code.
 
-[1]: <extensions/proposed/SYCL_EXT_ONEAPI_DEVICE_GLOBAL.asciidoc>
+[1]: <../extensions/proposed/SYCL_EXT_ONEAPI_DEVICE_GLOBAL.asciidoc>
 
 
 ## Requirements
@@ -403,7 +403,7 @@ has the `implement_in_csr` property).  See the
 [SPV\_INTEL\_global\_variable\_decorations][6] SPIR-V extension for details
 about all of these decorations.
 
-[6]: <extensions/DeviceGlobal/SPV_INTEL_global_variable_decorations.asciidoc>
+[6]: <../extensions/DeviceGlobal/SPV_INTEL_global_variable_decorations.asciidoc>
 
 The `sycl-post-link` tool also create a "SYCL/device globals" property set for
 each device code module that contains at least one device global variable.
@@ -614,7 +614,7 @@ The OpenCL backend has a proposed extension
 which can be easily used to implement these PI interfaces.  This DPC++ design
 depends upon implementation of that OpenCL extension.
 
-[10]: <extensions/DeviceGlobal/cl_intel_global_variable_access.asciidoc>
+[10]: <../extensions/DeviceGlobal/cl_intel_global_variable_access.asciidoc>
 
 The CUDA backend has existing APIs `cudaMemcpyToSymbol()` and
 `cudaMemcpyFromSymbol()` which can be used to implement these PI interfaces.

sycl/doc/extensions/C-CXX-StandardLibrary/DeviceLibExtensions.rst renamed to sycl/doc/design/DeviceLibExtensions.rst

@@ -48,7 +48,7 @@ Invoking `__devicelib_assert_read` after a kernel doesn't imply the kernel has
 assertion failed.
 
 See also: assert_extension_.
-.. _assert_extension: ../supported/SYCL_EXT_ONEAPI_ASSERT.asciidoc)
+.. _assert_extension: ../extensions/supported/SYCL_EXT_ONEAPI_ASSERT.asciidoc)
 
 cl_intel_devicelib_math
 ==========================