Commit Graph

150 Commits

Author SHA1 Message Date
Mateusz Jablonski
2bfdcb18d3 Refactor L0 ULT: move module fixture code to cpp file
Signed-off-by: Mateusz Jablonski <mateusz.jablonski@intel.com>
2022-11-25 17:38:07 +01:00
Dunajski, Bartosz
5266a95c03 Debug flag to control auto grf mode
Signed-off-by: Dunajski, Bartosz <bartosz.dunajski@intel.com>
2022-11-10 19:25:58 +01:00
Kacper Nowak
481cf36ad1 Add recompilation warning to debug messages
In case of recompilation from IR, apart from updating build log,
print debug message with recompilation warning.

Signed-off-by: Kacper Nowak <kacper.nowak@intel.com>
2022-11-09 14:19:54 +01:00
Luzynski, Sebastian Jozef
91c69e0fe7 Patch global variables and constants once
This change replaces mechanism of patching global constants and
variables in kernel per relocation to patching them only once. This
would improve linking time performance for kernels with multiple global
symbols.

Signed-off-by: Luzynski, Sebastian Jozef <sebastian.jozef.luzynski@intel.com>
2022-11-03 17:45:09 +01:00
Kacper Nowak
2285772c7e Disable zebin on pre-Gen12 platforms when debugger is active
On pre-Gen12 platform we use igfxdcd kernel module for debugging, which
does not support zebinary format.
- When platform is pre-Gen12 an and debugger is
attached, if binary format is zebin and it's not a builtin:
- If SPIR-V is available - force rebuild with zebin disabled
- Otherwise, return an error.
- Minor refactor: extend check for ir presence for each case of
rebuilt in OCL.

Related-To: NEO-7328
Signed-off-by: Kacper Nowak <kacper.nowak@intel.com>
2022-10-13 17:27:48 +02:00
Fabian Zwolinski
7953d15826 Print warning when kernel uses too much SLM
Instead of just returning proper error code in case of exceeding
available Shared Local Memory size we also want to print error message
to make debugging easier.

Related-To: NEO-7280
Signed-off-by: Fabian Zwolinski <fabian.zwolinski@intel.com>
2022-10-07 19:06:19 +02:00
Jaime Arteaga
f349343100 Enable l0_cache in L0
Name is different from the one used by opencl.

Signed-off-by: Jaime Arteaga <jaime.a.arteaga.molina@intel.com>
2022-10-07 07:01:54 +02:00
Fabian Zwolinski
1142404c0c Add error handling when kernel uses too much SLM
API Functions reporting error:
- clCreateKernel
- clEnqueueNDRangeKernel
- zeKernelCreate
- zeCommandListAppendLaunchKernel
- zeKernelSuggestGroupSize

Related-To: NEO-7280
Signed-off-by: Fabian Zwolinski <fabian.zwolinski@intel.com>
2022-10-06 16:45:42 +02:00
Neil R Spruit
8e2fcc9137 Allow init and execution of L0 without IGC unless it is needed
- Allow usage of L0 with IGC unless Debugger, Mid Thread Premption, or
  SPIRvs need to be compiled from modules. Native Binaries that are
already compiled for GENs will be usable.

Related-To: LOCI-3430

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-10-06 00:35:21 +02:00
Dunajski, Bartosz
ad2d3d0289 Remove not used method
Signed-off-by: Dunajski, Bartosz <bartosz.dunajski@intel.com>
2022-10-05 17:10:51 +02:00
Compute-Runtime-Validation
7aecea534f Revert "Default L0 Function & Global Symbols with fallback build for SPIRv"
This reverts commit 88b7a4f82d.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-09-23 07:07:04 +02:00
Dunajski, Bartosz
98db084b59 Debug flag to append api module build options
Signed-off-by: Dunajski, Bartosz <bartosz.dunajski@intel.com>
2022-09-22 16:03:58 +02:00
Dunajski, Bartosz
b2001bf265 L0: GRF mode debug flags support
Signed-off-by: Dunajski, Bartosz <bartosz.dunajski@intel.com>
2022-09-22 15:27:31 +02:00
Krystian Chmielewski
311b0b0020 Create input for linker during zebin decoding
Remove code duplication. Parsing zebin elf for relocations and symbols
is moved to decodeSingleDeviceBinary.

Signed-off-by: Krystian Chmielewski <krystian.chmielewski@intel.com>
2022-09-22 11:12:39 +02:00
Neil R Spruit
88b7a4f82d Default L0 Function & Global Symbols with fallback build for SPIRv
- Enabled default setting of Program & Global Symbols to be generated by
IGC when building L0 Modules with the ability to fallback to previous
behavior thru build failure checks.

- Enabled selective disable of default program or global symbol
generation thru debug variables.

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-09-22 02:40:51 +02:00
Zbigniew Zdanowicz
072686f283 Replace function term with kernel class specific names
Signed-off-by: Zbigniew Zdanowicz <zbigniew.zdanowicz@intel.com>
2022-09-12 12:00:24 +02:00
Dominik Dabek
7fa9cfd7fc Add missing cache policy isDebuggerActive values
Cache policy was not kept as WBP in some cases with debugger

Related-To: NEO-7003

Signed-off-by: Dominik Dabek <dominik.dabek@intel.com>
2022-09-09 10:38:41 +02:00
Kacper Nowak
710c8cf5ef Linker: add support for symbols with local binding
This commits add support for relocating
symbols with local binding and of functional type
(STB_LOCAL, STT_FUNC).

Related-To: NEO-7299
Signed-off-by: Kacper Nowak <kacper.nowak@intel.com>
2022-09-08 14:50:29 +02:00
Dunajski, Bartosz
16d9000429 Add option to change GRF mode
Signed-off-by: Dunajski, Bartosz <bartosz.dunajski@intel.com>
2022-09-06 13:00:17 +02:00
Compute-Runtime-Validation
c3fffcc8c0 Revert "Default L0 Function & Global Symbols with fallback build"
This reverts commit 2157f15594.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-09-03 07:08:52 +02:00
Neil R Spruit
2157f15594 Default L0 Function & Global Symbols with fallback build
- Enabled default setting of Program & Global Symbols to be generated by
IGC when building L0 Modules with the ability to fallback to previous
behavior thru build failure checks.

- Enabled selective disable of default program or global symbol
generation thru debug variables.

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-09-02 19:16:40 +02:00
Compute-Runtime-Validation
8bff54c773 Revert "Remove Support for reading a function pointer of a kernel"
This reverts commit 8d5792abf6.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-09-02 12:02:29 +02:00
Neil R Spruit
8d5792abf6 Remove Support for reading a function pointer of a kernel
Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-09-01 07:58:59 +02:00
Dominik Dabek
8cc0177f1c Change DG2 l1 cache policy to WB
With compiler LSC WAs this gives better performance.

If debugger is active, policy will not be changed ie.
will be WBP.

Related-To: NEO-7003

Signed-off-by: Dominik Dabek <dominik.dabek@intel.com>
2022-08-31 14:31:23 +02:00
Mateusz Hoppe
ee8af85341 L0Debug fixes for multi tile devices
- notify module allocation for (sub)device used to create module,
makeResident called within (sub)device contexts
- access ISA with vmHandle specific to device index

Related-To: NEO-5784

Signed-off-by: Mateusz Hoppe <mateusz.hoppe@intel.com>
2022-08-31 10:58:28 +02:00
Patryk Wrobel
de3e5d4d69 Improve files of SipKernel class
This change:
- removes redundant copying of std::vector
- removes unneeded includes from the header file
and introduces usage of forward declaration

Signed-off-by: Patryk Wrobel <patryk.wrobel@intel.com>
2022-08-29 15:18:11 +02:00
Compute-Runtime-Validation
2621460e80 Revert "Change DG2 l1 cache policy to WB"
This reverts commit a820e73dd7.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-08-27 08:04:19 +02:00
Dominik Dabek
a820e73dd7 Change DG2 l1 cache policy to WB
With compiler LSC WAs this gives better performance.

If debugger is active, policy will not be changed ie.
will be WBP.

Related-To: NEO-7003

Signed-off-by: Dominik Dabek <dominik.dabek@intel.com>
2022-08-26 12:58:45 +02:00
Igor Venevtsev
d40173e47a L0Debug Win: Always generate module create/destroy events
We have to always generate module create/destroy events to give
debugger a chance to insert bp, read/write debuggee memory  etc
even there is no debug data was generated. In this case ELF will not be
reported to debugger, just ISA GpuVA

Related-To: NEO-6723

Signed-off-by: Igor Venevtsev <igor.venevtsev@intel.com>
2022-08-24 14:58:30 +02:00
Mateusz Jablonski
7a586111f2 Don't program implicit args for builtin kernels
Related-To: HSD-18023305082

Signed-off-by: Mateusz Jablonski <mateusz.jablonski@intel.com>
2022-08-18 20:10:10 +02:00
Neil R Spruit
0101e80b00 Verify L0 Image Format Support for SPIRv
- Check that the Image Format is valid for the image
argument to a SPIRv module. If the Image is invalid
return ZE_RESULT_ERROR_UNSUPPORTED_IMAGE_FORMAT.

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-08-12 17:57:17 +02:00
Compute-Runtime-Validation
f4c8a146eb Revert "Remove Support for reading a function pointer of a kernel"
This reverts commit 299dcb8bd2.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-08-02 11:58:31 +02:00
Neil R Spruit
299dcb8bd2 Remove Support for reading a function pointer of a kernel
Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-08-01 12:28:39 +02:00
Igor Venevtsev
4cb9ad5d55 Debugger L0 Win: implement module destroy event
Related-To: NEO-6723

Signed-off-by: Igor Venevtsev <igor.venevtsev@intel.com>
2022-07-28 15:55:40 +02:00
Igor Venevtsev
d79e799bea Debugger L0 win: implement zetAcknowledgeEvent API
Related-To: NEO-6723

Signed-off-by: Igor Venevtsev <igor.venevtsev@intel.com>
2022-07-22 16:46:27 +02:00
Igor Venevtsev
547dd59272 Debugger L0 Win: Implement CREATE_DEBUG_DATA/MODULE_CREATE events handling
Related-To: NEO-6723

Signed-off-by: Igor Venevtsev <igor.venevtsev@intel.com>
2022-07-20 13:20:50 +02:00
Tratnack, Geoffrey
16bd39650f Adding Implementation of zetKernelGetProfileInfo
Signed-off-by: Tratnack, Geoffrey <geoffrey.tratnack@intel.com>
2022-07-07 21:45:08 +02:00
Mateusz Hoppe
5956aea18d Limit header includes from level_zero device.h
- remove including debugger_l0.h from device.h
- add getL0Debugger() to shared NEO Device

Signed-off-by: Mateusz Hoppe <mateusz.hoppe@intel.com>
2022-07-06 16:41:17 +02:00
Szymon Morek
9a26e505a8 Pass LSC policy to IGC in build options
Related-To: NEO-7003


Signed-off-by: Szymon Morek <szymon.morek@intel.com>
2022-06-21 11:51:56 +02:00
Compute-Runtime-Validation
3dd6e3f125 Revert "Enable Symbol Table Generation by Default for L0 modules"
This reverts commit c2e3d24a1d.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-06-15 10:28:47 +02:00
Compute-Runtime-Validation
930db249ff Revert "Enable Global Symbol Generation by Default for L0 modules"
This reverts commit 9996228281.

Signed-off-by: Compute-Runtime-Validation <compute-runtime-validation@intel.com>
2022-06-15 08:45:53 +02:00
Neil R Spruit
9996228281 Enable Global Symbol Generation by Default for L0 modules
Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-06-14 22:36:08 +02:00
Neil R Spruit
e2132de30e L0 Function Pointer Handling for kernels & correct error
- Enable Support for L0 Function Pointers for Kernels and Exported
Functions
- Return ZE_RESULT_ERROR_INVALID_FUNCTION_NAME given no function of that
name is a kernel or exported function

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-06-14 18:19:01 +02:00
Neil R Spruit
c2e3d24a1d Enable Symbol Table Generation by Default for L0 modules
- Added EnableProgramSymbolTableGeneration to enable or disable default
behavior for IGC to generate the program symbol tables for L0 modules
with exported functions.
- Default value set to true to add -library-compilation to all module
builds.

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-06-14 18:10:24 +02:00
Bartosz Dunajski
5510dc7daa Add adjustHwInfoForIgc support
Signed-off-by: Bartosz Dunajski <bartosz.dunajski@intel.com>
2022-06-13 14:31:53 +02:00
Mateusz Jablonski
747c2d3d83 Cleanup includes in cmdlist, cmdqueue, context, device, driver and module
Signed-off-by: Mateusz Jablonski <mateusz.jablonski@intel.com>
2022-06-03 02:54:41 +02:00
Mateusz Jablonski
16fd791a7c Fix L0: reuse partially linked segments when performing dynamic linking
When module is created but not fully linked we don't copy ISA to GPU
however we need to store patched segments for dynamic linking

Related-To HSD-18022114161

Signed-off-by: Mateusz Jablonski <mateusz.jablonski@intel.com>
2022-05-30 17:12:30 +02:00
Neil R Spruit
8c894fad5e Change Packed binary Returned in L0 GetNative to be the Target Device binary
- Packed Binary Returned to a customer in L0 GetNative is now
the target device's native binary.

Signed-off-by: Neil R Spruit <neil.r.spruit@intel.com>
2022-05-25 21:27:45 +02:00
Katarzyna Cencelewska
96e1eb7467 Move variables baseDieRev and baseDieA0Masked from xe_hpc to pvc
Pvc specific variables should be located in pvc struct

Related-To: NEO-6738
Signed-off-by: Katarzyna Cencelewska <katarzyna.cencelewska@intel.com>
2022-05-17 12:19:16 +02:00
Kamil Kopryk
99db73c034 Add debug flag to fail build program with stateful access
I've added debug flag FailBuildProgramWithStatefulAccess which makes
possible to fail build program/module creation
with stateful access(except builtins) on
pvc and later platforms.

Related-To: NEO-6075
Signed-off-by: Kamil Kopryk <kamil.kopryk@intel.com>
2022-05-17 11:01:05 +02:00