diff --git a/src/intel/vulkan/anv_physical_device.c b/src/intel/vulkan/anv_physical_device.c index cabfa63d395..3dc9cfc8e8e 100644 --- a/src/intel/vulkan/anv_physical_device.c +++ b/src/intel/vulkan/anv_physical_device.c @@ -6,6 +6,7 @@ #include "anv_api_version.h" #include "anv_measure.h" +#include "dev/intel_debug.h" #include "i915/anv_device.h" #include "xe/anv_device.h" @@ -133,7 +134,8 @@ static void get_device_extensions(const struct anv_physical_device *device, struct vk_device_extension_table *ext) { - const bool rt_enabled = ANV_SUPPORT_RT && device->info.has_ray_tracing; + const bool rt_enabled = ANV_SUPPORT_RT && device->info.has_ray_tracing && + !intel_use_jay_any_stage(&device->info); const bool hw_video_encode_supported = device->info.verx10 < 125; const bool video_encode_enabled = hw_video_encode_supported && (device->instance->debug & ANV_DEBUG_VIDEO_ENCODE); @@ -2779,7 +2781,8 @@ anv_physical_device_try_create(struct vk_instance *vk_instance, device->has_cooperative_matrix = (device->info.has_systolic || debug_get_bool_option("INTEL_LOWER_DPAS", false)) && - device->info.cooperative_matrix_configurations[0].scope != INTEL_CMAT_SCOPE_NONE; + device->info.cooperative_matrix_configurations[0].scope != INTEL_CMAT_SCOPE_NONE && + !intel_use_jay_any_stage(&device->info); if (is_virtio) { struct util_sync_provider *sync = intel_virtio_sync_provider(fd); diff --git a/src/intel/vulkan/anv_shader_compile.c b/src/intel/vulkan/anv_shader_compile.c index 907f0162e19..63e20dd901d 100644 --- a/src/intel/vulkan/anv_shader_compile.c +++ b/src/intel/vulkan/anv_shader_compile.c @@ -15,6 +15,7 @@ #include "compiler/brw/brw_nir.h" #include "compiler/brw/brw_nir_rt.h" #include "compiler/intel_nir.h" +#include "compiler/jay/jay.h" #include "git_sha1.h" @@ -833,6 +834,7 @@ anv_shader_compile_vs(struct anv_device *device, char **error_str) { const struct brw_compiler *compiler = device->physical->compiler; + const struct intel_device_info *devinfo = compiler->devinfo; nir_shader *nir = shader_data->info->nir; shader_data->num_stats = 1; @@ -850,7 +852,17 @@ anv_shader_compile_vs(struct anv_device *device, .prog_data = &shader_data->prog_data.vs, }; - shader_data->code = (void *)brw_compile_vs(compiler, ¶ms); + if (intel_use_jay(devinfo, nir->info.stage)) { + struct jay_shader_bin *bin = + jay_compile(devinfo, mem_ctx, nir, + (union brw_any_prog_data *) params.prog_data, + (union brw_any_prog_key *) params.key); + + shader_data->code = (void *) bin->kernel; + } else { + shader_data->code = (void *) brw_compile_vs(compiler, ¶ms); + } + *error_str = params.base.error_str; } @@ -1040,6 +1052,7 @@ anv_shader_compile_fs(struct anv_device *device, char **error_str) { const struct brw_compiler *compiler = device->physical->compiler; + const struct intel_device_info *devinfo = compiler->devinfo; nir_shader *nir = shader_data->info->nir; /* When using Primitive Replication for multiview, each view gets its own @@ -1072,7 +1085,17 @@ anv_shader_compile_fs(struct anv_device *device, .max_polygons = UCHAR_MAX, }; - shader_data->code = (void *)brw_compile_fs(compiler, ¶ms); + if (intel_use_jay(devinfo, nir->info.stage)) { + struct jay_shader_bin *bin = + jay_compile(devinfo, mem_ctx, nir, + (union brw_any_prog_data *) params.prog_data, + (union brw_any_prog_key *) params.key); + + shader_data->code = (void *) bin->kernel; + } else { + shader_data->code = (void *) brw_compile_fs(compiler, ¶ms); + } + *error_str = params.base.error_str; shader_data->num_stats = (uint32_t)!!shader_data->prog_data.fs.dispatch_multi + @@ -1101,6 +1124,7 @@ anv_shader_compile_cs(struct anv_device *device, char **error_str) { const struct brw_compiler *compiler = device->physical->compiler; + const struct intel_device_info *devinfo = compiler->devinfo; nir_shader *nir = shader_data->info->nir; shader_data->num_stats = 1; @@ -1118,7 +1142,21 @@ anv_shader_compile_cs(struct anv_device *device, .prog_data = &shader_data->prog_data.cs, }; - shader_data->code = (void *)brw_compile_cs(compiler, ¶ms); + if (intel_use_jay(devinfo, nir->info.stage)) { + struct jay_shader_bin *bin = jay_compile(devinfo, mem_ctx, nir, + (union brw_any_prog_data*)params.prog_data, + (union brw_any_prog_key*)params.key); + + shader_data->code = (void*)bin->kernel; + shader_data->stats[0] = bin->stats; + + params.prog_data->local_size[0] = nir->info.workgroup_size[0]; + params.prog_data->local_size[1] = nir->info.workgroup_size[1]; + params.prog_data->local_size[2] = nir->info.workgroup_size[2]; + } else { + shader_data->code = (void*)brw_compile_cs(compiler, ¶ms); + } + *error_str = params.base.error_str; } diff --git a/src/intel/vulkan/meson.build b/src/intel/vulkan/meson.build index 1f44c2dec95..22ed0712599 100644 --- a/src/intel/vulkan/meson.build +++ b/src/intel/vulkan/meson.build @@ -277,7 +277,8 @@ libvulkan_intel = shared_library( idep_nir, idep_genxml, idep_vulkan_util, idep_vulkan_wsi, idep_vulkan_runtime, idep_mesautil, idep_xmlconfig, idep_intel_driver_ds, idep_intel_dev, idep_intel_blorp, - idep_intel_compiler_brw, idep_libvulkan_intel_decoder + idep_intel_compiler_brw, idep_libvulkan_intel_decoder, + idep_intel_compiler_jay, ], c_args : anv_flags, gnu_symbol_visibility : 'hidden', @@ -315,7 +316,7 @@ if with_tests dep_thread, dep_dl, dep_m, anv_deps, idep_nir, idep_vulkan_util, idep_vulkan_wsi, idep_vulkan_runtime, idep_mesautil, idep_intel_dev, idep_intel_blorp, - idep_intel_compiler_brw, idep_libvulkan_intel_decoder, + idep_intel_compiler_brw, idep_intel_compiler_jay, idep_libvulkan_intel_decoder, ], c_args : anv_flags, gnu_symbol_visibility : 'hidden',