From 7e6296bec2521e2b398fa796e006fbd22fa3a46d Mon Sep 17 00:00:00 2001 From: Pavel V Chupin Date: Wed, 10 Mar 2021 20:17:49 -0800 Subject: [PATCH 1/6] [SYCL][Doc] Update compiler options descriptions * Copied renewed option descriptions from other sources * Added few missing options * Added "auto" variant for -fsycl-device-code-split * Marked some of the options as [DEPRECATED] which are not expected to be used at this point. * Marked some of the options as [EXPERIMENTAL] which are expected to be used only in limitted cases. * Updated formating and sections structure * Removed fat static library case as it's not expected to be used now through special option. Instead fat static libraries should be just added to the command line normally as any static library. Signed-off-by: Pavel V Chupin --- sycl/doc/UsersManual.md | 247 +++++++++++++++++++++------------------- 1 file changed, 128 insertions(+), 119 deletions(-) diff --git a/sycl/doc/UsersManual.md b/sycl/doc/UsersManual.md index 8a0ac1be8c725..26aaffa5e0d22 100644 --- a/sycl/doc/UsersManual.md +++ b/sycl/doc/UsersManual.md @@ -1,55 +1,108 @@ # Users Manual -The DPC++ Compiler contains many options to generate the desired binaries for -your application. +This is the list of SYCL specific options supported by compiler and some +examples. -## SYCL specific command line options +## Generic options **`-fsycl`** - General enabling option for SYCL compilation mode. This option enables - offload compilation for a given target using the `-fsycl-targets` option. - When the `-fsycl-targets` option is not provided, the default triple is - `spir64-unknown-unknown-sycldevice`. + General enabling option for SYCL compilation and linking mode. List of + targets can be specified with `-fsycl-targets`. This is fundamental option + for any SYCL compilation. All other SYCL specific options require it. -**`-fsycl-targets=`** +**`-fsycl-targets=[,...,]`** - A comma separated list of triples to specify the device target(s) to - generate code for. This option is only valid when used with `-fsycl`. + Tells the compiler to generate code for specified devices. T is a target + triple device name. You can specify more than one target, comma separated. + The following triples are supported: + * spir64-unknown-unknown-sycldevice - this is the default generic SPIR-V + target; + * spir64_x86_64-unknown-unknown-sycldevice - generate code ahead of time + for x86_64 CPUs; + * spir64_fpga-unknown-unknown-sycldevice - generate code ahead of time for + Intel FPGA; + * spir64_gen-unknown-unknown-sycldevice - generate code ahead of time for + Intel Processor Graphics; + * nvptx64-nvidia-cuda-sycldevice - generate code ahead of time for CUDA + target; -### Target toolchain options. +## Language options -**`-Xsycl-target-backend= `** +**`-sycl-std=`** [EXPERIMENTAL] - Pass to the SYCL based backend identified by . + SYCL language standard to compile for. Possible values: + * 121 - SYCL 1.2.1 + * 2020 - SYCL 2020 + It doesn't guarantee specific standard compliance, but some selected + compiler features change behavior. + It is under development and not recommended to use in production + environment. + Default value is 2020. -**`-Xsycl-target-backend `** +**`-f[no-]sycl-unnamed-lambda`** - Pass to the SYCL based target backend. + Enables/Disables unnamed SYCL lambda kernels support. + Enabled by default. -**`-Xsycl-target-frontend= `** +## Optimization options - Pass to the SYCL based target frontend identified by . +**`-f[no-]sycl-early-optimizations`** -**`-Xsycl-target-frontend `** + Enables (or disables) intermediate representation optimization pipeline + before translation to SPIR-V. Have effect only if optimizations are turned + on by standard compiler options (-O1 or higher). + Enabled by default. - Pass to the SYCL based target frontend. +**`-f[no-]sycl-dead-args-optimization`** -**`-Xsycl-target-linker= `** + Enables (or disables) LLVM IR dead argument elimination pass to remove + unused arguments for the kernel functions before translation to SPIR-V. + Disabled by default. - Pass to the SYCL based target linker identified by . +**`-f[no-]sycl-id-queries-fit-in-int`** -**`-Xsycl-target-linker `** + Assume/Do not assume that SYCL ID queries fit within MAX_INT. It assumes + that these values fit within MAX_INT: + * id class get() member function and operator[] + * item class get_id() member function and operator[] + * nd_item class get_global_id()/get_global_linear_id() member functions + Enabled by default. - Pass to the SYCL based target linker. +## Target toolchain options -### Link options +**`-Xsycl-target-backend= "options"`** +**`-Xs "options"`** + + Pass "options" to the backend of target device compiler, specified by + triple T. The backend of device compiler generates target machine code from + intermediate representation. This option can be used to tune code + generation for a specific target. For offline compilation it happens during + compilation. For online compilation options are saved in a fat binary to be + used later, during runtime compilation. + -Xs is a shortcut to pass options to all backends specified (or default + one). + +**`-Xsycl-target-frontend= "options"`** + + Pass "options" to the frontend of target device compiler, specified by + triple T. This option can be used to control of intermediate representation + generation during offline or online compilation. + +**`-Xsycl-target-linker= "options"`** + + Pass "options" to the device code linker, when linking multiple device + object modules. T is specific target device triple. + +## Link options **`-fsycl-link`** - Generate partially linked device object to be used with the host link. + Link device object modules and wrap those into a host-compatible object + module that can be linked later by any standard host linker into the final + fat binary. -**`-fsycl-link-targets=`** +**`-fsycl-link-targets=`** [DEPRECATED] Specify comma-separated list of triples SYCL offloading targets to produce linked device images. Used in a link step to link device code for given @@ -57,7 +110,7 @@ your application. of the common prefix taken from the -o option and the triple string. Does not produce fat binary and must be used together with -fsycl. -**`-fsycl-add-targets=`** +**`-fsycl-add-targets=`** [DEPRECATED] Add arbitrary device images to the fat binary being linked @@ -68,7 +121,7 @@ your application. image for the target triple it is paired with, and offload bundler is invoked to do the actual bundling. -**`-foffload-static-lib=`** +**`-foffload-static-lib=`** [DEPRECATED] Link with fat static library. @@ -81,95 +134,84 @@ your application. specified with `-foffload-static-lib` are treated as host libraries and are only used during the final host link. -**`-foffload-whole-static-lib=`** +**`-foffload-whole-static-lib=`** [DEPRECATED] Similar to `-foffload-static-lib` but uses the whole archive when performing the device code extraction. This is helpful when creating shared objects from fat static archives. -**`-fsycl-device-code-split=`** +**`-fsycl-device-code-split=`** - Perform SYCL device code split. There are three possible values for this - option: - - per_kernel - a separate device code module is created for each SYCL - kernel. Each device code module will contain a kernel and all its - dependencies, i.e. called functions and used variables. - - per_source - a separate device code module is created for each source - (translation unit). Each device code module will contain a bunch of kernels - grouped on per-source basis and all their dependencies, i.e. all used - variables and called functions, including the `SYCL_EXTERNAL` macro-marked - functions from other translation units. - - off - no device code split. - NOTE: By default device code split is 'off' - all kernels go into a - single module. + Specifies SYCL device code module assembly. Mode is one of the following: + * per_kernel - creates a separate device code module for each SYCL kernel. + Each device code module will contain a kernel and all its dependencies, + such as called functions and used variables. + * per_source - creates a separate device code module for each source + (translation unit). Each device code module will contain a bunch of + kernels grouped on per-source basis and all their dependencies, such as + all used variables and called functions, including the `SYCL_EXTERNAL` + macro-marked functions from other translation units. + * off - creates a single module for all kernels. + * auto - the compiler will use a heuristic to select the best way of + splitting device code. This is default mode. -**`-fsycl-device-code-split`** +**`-f[no-]sycl-device-lib=[,,...]`** - Perform SYCL device code split in the per_source mode, i.e. create a - separate device code module for each source (translation unit). + Enables/disables linking of the device libraries. Supported libraries: + libm-fp32, libm-fp64, libc, all. Use of 'all' will enable/disable all of + the device libraries. -### Intel FPGA specific options +## Intel FPGA specific options **`-fintelfpga`** - Perform ahead of time compilation for Intel FPGA, which relies on the - external tool `aoc` being available in the `PATH`. - - This option is roughly equivalent to - `-fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -g -MMD -lOpenCL`. - - It is incompatible with `-fsycl-targets=...`; if ahead of time compilation - is needed for multiple backends (e.g. Intel FPGA, Intel GPU, etc.), the - alternative form based on - `-fsycl-targets=spir64_fpga-unknown-unknown-sycldevice` should be used - instead. + Perform ahead of time compilation for Intel FPGA. It sets the target to + FPGA and turns on the debug options that are needed to generate FPGA + reports. It is functionally equivalent shortcut to + `-fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -g -MMD` on Linux + and `-fsycl-targets=spir64_fpga-unknown-unknown-sycldevice -Zi -MMD` on + Windows. -**`-fsycl-link=`** +**`-fsycl-link=`** - Generate partially linked device and host object to be used at various - stages of compilation. Takes the device binary(s) generated from a `-fsycl` - enabled compilation and wrap to create a host linkable object. This option - is enabled only in ahead of time compilation mode fore FPGA (i.e. when - `-fintelfpga` is set). + Controls FPGA target binary output format. Same as -fsycl-link, but + optional output can be one of the following: + * early - generate html reports and an intermediate object file that avoids + a full Quartus compile. Usually takes minutes to generate. Link can later + be resumed from this point using -fsycl-link=image. + * image - generate a bitstream which is ready to be linked and used on a + FPGA board. Usually takes hours to generate. **`-reuse-exe=`** - Speed up FPGA aoc compile if the device code in is unchanged. + Speed up FPGA backend compilation if the device code in is + unchanged. If it's safe to do so the compiler will re-use the device binary + embedded within it. This can be used to minimize or avoid long Quartus + compile times for FPGA targets when the device code is unchanged. -### Other options +## Other options **`-fsycl-device-only`** - Compile only SYCL device code. + Compile only device part of the code and discard host part. -**`-fsycl-use-bitcode`** +**`-fsycl-use-bitcode`** [DEPRECATED] Emit SYCL device code in LLVM-IR bitcode format. When disabled, SPIR-V is emitted. Default is true. -**`-fno-sycl-use-bitcode`** +**`-fno-sycl-use-bitcode`** [DEPRECATED] Use SPIR-V instead of LLVM bitcode in fat objects. -**`-sycl-std=`** +**`-fsycl-help[=backend]`** - SYCL language standard to compile for. + Emit help information from device compiler backend. Backend can be one of + the following: "x86_64", "fpga", "gen", or "all". Specifying "all" is the + same as specifying -fsycl-help with no argument and emits help for all + backends. -**`-fsycl-help`** - - Emit help information from all of the offline compilation tools. - -**`-fsycl-help=`** - - Emit help information from the offline compilation tool associated with the - given architecture argument. Supported architectures: `x86_64`, `fpga` and - `gen`. - -**`-fsycl-unnamed-lambda`** - - Allow unnamed SYCL lambda kernels. - -## SYCL device code compilation +# Example: SYCL device code compilation To invoke SYCL device compiler set `-fsycl-device-only` flag. @@ -184,36 +226,3 @@ By default the output format for SYCL device is LLVM bytecode. ```console $ clang++ -fsycl-device-only -fno-sycl-use-bitcode sycl-app.cpp -o sycl-app.spv ``` - -## Static archives with SYCL device code - -The DPC++ Compiler contains support to create and use static archives that -contain device enabled fat objects. - -### Build your objects - -```console -$ clang++ -fsycl sycl-app1.cpp sycl-app2.cpp -c -``` - -### Create the static archive - -Build the static archive in the same manner as you would any other normal -static archive, using the objects that were created using the above step. - -```console -$ ar cr libsyclapp.a sycl-app1.o sycl-app2.o -``` - -### Use the static archive - -Once you have created the archive, you can use it when creating your final -application. The fat archives are treated differently than a regular archive -so the option `-foffload-static-lib` is used to signify the needed behavior. - -```console -$ clang++ -fsycl sycl-main.cpp -foffload-static-lib=libsyclapp.a -``` - -Use of `-foffload-static-lib` is required or the library will be treated as -a normal archive. From 704fe0b62505ddb25b3dba7213f7b031b8f80901 Mon Sep 17 00:00:00 2001 From: Pavel V Chupin Date: Thu, 11 Mar 2021 23:18:15 -0800 Subject: [PATCH 2/6] Applying review comments --- sycl/doc/UsersManual.md | 9 +++++++-- 1 file changed, 7 insertions(+), 2 deletions(-) diff --git a/sycl/doc/UsersManual.md b/sycl/doc/UsersManual.md index 26aaffa5e0d22..bf2b32e46ae56 100644 --- a/sycl/doc/UsersManual.md +++ b/sycl/doc/UsersManual.md @@ -3,6 +3,10 @@ This is the list of SYCL specific options supported by compiler and some examples. +Options marked as [DEPRECATED] are going to be removed in some future updates. +Options marked as [EXPERIMENTAL] are expected to be used only in limitted cases +and not recommended to use in production environment. + ## Generic options **`-fsycl`** @@ -15,7 +19,7 @@ examples. Tells the compiler to generate code for specified devices. T is a target triple device name. You can specify more than one target, comma separated. - The following triples are supported: + The following triples are supported by default: * spir64-unknown-unknown-sycldevice - this is the default generic SPIR-V target; * spir64_x86_64-unknown-unknown-sycldevice - generate code ahead of time @@ -24,6 +28,7 @@ examples. Intel FPGA; * spir64_gen-unknown-unknown-sycldevice - generate code ahead of time for Intel Processor Graphics; + Available in special build configuration: * nvptx64-nvidia-cuda-sycldevice - generate code ahead of time for CUDA target; @@ -193,7 +198,7 @@ examples. **`-fsycl-device-only`** - Compile only device part of the code and discard host part. + Compile only device part of the code and ignore host part. **`-fsycl-use-bitcode`** [DEPRECATED] From 14e6df65d3be3f62d8ce3f622c9554ecf1b591bd Mon Sep 17 00:00:00 2001 From: Pavel Chupin <45979248+pvchupin@users.noreply.github.com> Date: Sat, 13 Mar 2021 16:41:03 -0800 Subject: [PATCH 3/6] Update sycl/doc/UsersManual.md Co-authored-by: mdtoguchi <47896532+mdtoguchi@users.noreply.github.com> --- sycl/doc/UsersManual.md | 3 ++- 1 file changed, 2 insertions(+), 1 deletion(-) diff --git a/sycl/doc/UsersManual.md b/sycl/doc/UsersManual.md index bf2b32e46ae56..c4e968da43350 100644 --- a/sycl/doc/UsersManual.md +++ b/sycl/doc/UsersManual.md @@ -85,7 +85,8 @@ and not recommended to use in production environment. generation for a specific target. For offline compilation it happens during compilation. For online compilation options are saved in a fat binary to be used later, during runtime compilation. - -Xs is a shortcut to pass options to all backends specified (or default + -Xs is a shortcut to pass "options" to all backends specified via the + '-fsycl-targets' option (or default one). **`-Xsycl-target-frontend= "options"`** From e53afe4576ca8dd423283adc8c50e663f37a368f Mon Sep 17 00:00:00 2001 From: Pavel Chupin <45979248+pvchupin@users.noreply.github.com> Date: Sat, 13 Mar 2021 16:42:50 -0800 Subject: [PATCH 4/6] Update sycl/doc/UsersManual.md Co-authored-by: mdtoguchi <47896532+mdtoguchi@users.noreply.github.com> --- sycl/doc/UsersManual.md | 6 +++--- 1 file changed, 3 insertions(+), 3 deletions(-) diff --git a/sycl/doc/UsersManual.md b/sycl/doc/UsersManual.md index c4e968da43350..de35b2e0f5963 100644 --- a/sycl/doc/UsersManual.md +++ b/sycl/doc/UsersManual.md @@ -82,9 +82,9 @@ and not recommended to use in production environment. Pass "options" to the backend of target device compiler, specified by triple T. The backend of device compiler generates target machine code from intermediate representation. This option can be used to tune code - generation for a specific target. For offline compilation it happens during - compilation. For online compilation options are saved in a fat binary to be - used later, during runtime compilation. + generation for a specific target. The options are used during offline compilation + and are also saved in a fat binary when used for online compilation to be used + at runtime compilation. -Xs is a shortcut to pass "options" to all backends specified via the '-fsycl-targets' option (or default one). From 3d68fa1f4f37c87a4a065cbbb2280653e0fca2d5 Mon Sep 17 00:00:00 2001 From: Pavel V Chupin Date: Sat, 13 Mar 2021 16:57:05 -0800 Subject: [PATCH 5/6] Applying review comments --- sycl/doc/UsersManual.md | 19 +++++++------------ 1 file changed, 7 insertions(+), 12 deletions(-) diff --git a/sycl/doc/UsersManual.md b/sycl/doc/UsersManual.md index de35b2e0f5963..c58beaba5aae0 100644 --- a/sycl/doc/UsersManual.md +++ b/sycl/doc/UsersManual.md @@ -48,7 +48,7 @@ and not recommended to use in production environment. **`-f[no-]sycl-unnamed-lambda`** Enables/Disables unnamed SYCL lambda kernels support. - Enabled by default. + Disabled by default. ## Optimization options @@ -82,12 +82,10 @@ and not recommended to use in production environment. Pass "options" to the backend of target device compiler, specified by triple T. The backend of device compiler generates target machine code from intermediate representation. This option can be used to tune code - generation for a specific target. The options are used during offline compilation - and are also saved in a fat binary when used for online compilation to be used - at runtime compilation. + generation for a specific target. The "options" are used during offline + compilation and are also saved in a fat binary for online compilation. -Xs is a shortcut to pass "options" to all backends specified via the - '-fsycl-targets' option (or default - one). + '-fsycl-targets' option (or default one). **`-Xsycl-target-frontend= "options"`** @@ -201,14 +199,11 @@ and not recommended to use in production environment. Compile only device part of the code and ignore host part. -**`-fsycl-use-bitcode`** [DEPRECATED] +**`-f[no-]sycl-use-bitcode`** Emit SYCL device code in LLVM-IR bitcode format. When disabled, SPIR-V is - emitted. Default is true. - -**`-fno-sycl-use-bitcode`** [DEPRECATED] - - Use SPIR-V instead of LLVM bitcode in fat objects. + emitted. + Enabled by default. **`-fsycl-help[=backend]`** From f8189215c608a81e165186fe206cf4341d94aa64 Mon Sep 17 00:00:00 2001 From: Pavel V Chupin Date: Tue, 16 Mar 2021 20:28:20 -0700 Subject: [PATCH 6/6] Applying review comments --- sycl/doc/UsersManual.md | 15 ++++++++++----- 1 file changed, 10 insertions(+), 5 deletions(-) diff --git a/sycl/doc/UsersManual.md b/sycl/doc/UsersManual.md index c58beaba5aae0..a8ba4715ff645 100644 --- a/sycl/doc/UsersManual.md +++ b/sycl/doc/UsersManual.md @@ -17,8 +17,11 @@ and not recommended to use in production environment. **`-fsycl-targets=[,...,]`** - Tells the compiler to generate code for specified devices. T is a target - triple device name. You can specify more than one target, comma separated. + Enables ahead of time (AOT) compilation for specified device targets. T is + a compiler target triple string, representing a target device architecture. + You can specify more than one target, comma separated. Default just in time + (JIT) compilation target can be added to the list to produce a combination + of AOT and JIT code in the resulting fat binary. The following triples are supported by default: * spir64-unknown-unknown-sycldevice - this is the default generic SPIR-V target; @@ -63,6 +66,7 @@ and not recommended to use in production environment. Enables (or disables) LLVM IR dead argument elimination pass to remove unused arguments for the kernel functions before translation to SPIR-V. + Currently has effect only on spir64\* targets. Disabled by default. **`-f[no-]sycl-id-queries-fit-in-int`** @@ -82,8 +86,9 @@ and not recommended to use in production environment. Pass "options" to the backend of target device compiler, specified by triple T. The backend of device compiler generates target machine code from intermediate representation. This option can be used to tune code - generation for a specific target. The "options" are used during offline - compilation and are also saved in a fat binary for online compilation. + generation for a specific target. The "options" are used during AOT + compilation. For JIT compilation "options" are saved in a fat binary and + used when code is JITed during runtime. -Xs is a shortcut to pass "options" to all backends specified via the '-fsycl-targets' option (or default one). @@ -199,7 +204,7 @@ and not recommended to use in production environment. Compile only device part of the code and ignore host part. -**`-f[no-]sycl-use-bitcode`** +**`-f[no-]sycl-use-bitcode`** [EXPERIMENTAL] Emit SYCL device code in LLVM-IR bitcode format. When disabled, SPIR-V is emitted.