Compare commits
12 Commits
master
...
dxil-offse
Author | SHA1 | Date |
---|---|---|
Hans-Kristian Arntzen | d8e31d9b54 | |
Philip Rebohle | ac1c94022a | |
Philip Rebohle | 84d45a39f0 | |
Philip Rebohle | 090db312dc | |
Philip Rebohle | 3d2ae44737 | |
Philip Rebohle | e0f1675838 | |
Philip Rebohle | 95f6c8db69 | |
Philip Rebohle | 98917ec6ed | |
Philip Rebohle | 2021de2738 | |
Philip Rebohle | 1d64b5c4b1 | |
Philip Rebohle | 1c35cee200 | |
Philip Rebohle | 05dcb7ac16 |
|
@ -50,14 +50,9 @@ struct hash_map
|
||||||
uint32_t used_count;
|
uint32_t used_count;
|
||||||
};
|
};
|
||||||
|
|
||||||
static inline void *hash_map_ptr_offset(void *ptr, size_t offset)
|
|
||||||
{
|
|
||||||
return ((char*)ptr) + offset;
|
|
||||||
}
|
|
||||||
|
|
||||||
static inline struct hash_map_entry *hash_map_get_entry(struct hash_map *hash_map, uint32_t entry_idx)
|
static inline struct hash_map_entry *hash_map_get_entry(struct hash_map *hash_map, uint32_t entry_idx)
|
||||||
{
|
{
|
||||||
return hash_map_ptr_offset(hash_map->entries, hash_map->entry_size * entry_idx);
|
return void_ptr_offset(hash_map->entries, hash_map->entry_size * entry_idx);
|
||||||
}
|
}
|
||||||
|
|
||||||
static inline uint32_t hash_map_get_entry_idx(struct hash_map *hash_map, uint32_t hash_value)
|
static inline uint32_t hash_map_get_entry_idx(struct hash_map *hash_map, uint32_t hash_value)
|
||||||
|
@ -98,7 +93,7 @@ static inline bool hash_map_grow(struct hash_map *hash_map)
|
||||||
for (i = 0; i < old_count; i++)
|
for (i = 0; i < old_count; i++)
|
||||||
{
|
{
|
||||||
/* Relocate existing entries one by one */
|
/* Relocate existing entries one by one */
|
||||||
struct hash_map_entry *old_entry = hash_map_ptr_offset(old_entries, i * hash_map->entry_size);
|
struct hash_map_entry *old_entry = void_ptr_offset(old_entries, i * hash_map->entry_size);
|
||||||
|
|
||||||
if (old_entry->flags & HASH_MAP_ENTRY_OCCUPIED)
|
if (old_entry->flags & HASH_MAP_ENTRY_OCCUPIED)
|
||||||
{
|
{
|
||||||
|
|
|
@ -271,4 +271,9 @@ static inline size_t vkd3d_wcslen(const WCHAR *wstr, size_t wchar_size)
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
|
static inline void *void_ptr_offset(void *ptr, size_t offset)
|
||||||
|
{
|
||||||
|
return ((char*)ptr) + offset;
|
||||||
|
}
|
||||||
|
|
||||||
#endif /* __VKD3D_COMMON_H */
|
#endif /* __VKD3D_COMMON_H */
|
||||||
|
|
|
@ -187,6 +187,7 @@ enum vkd3d_shader_interface_flag
|
||||||
{
|
{
|
||||||
VKD3D_SHADER_INTERFACE_PUSH_CONSTANTS_AS_UNIFORM_BUFFER = 0x00000001u,
|
VKD3D_SHADER_INTERFACE_PUSH_CONSTANTS_AS_UNIFORM_BUFFER = 0x00000001u,
|
||||||
VKD3D_SHADER_INTERFACE_BINDLESS_CBV_AS_STORAGE_BUFFER = 0x00000002u,
|
VKD3D_SHADER_INTERFACE_BINDLESS_CBV_AS_STORAGE_BUFFER = 0x00000002u,
|
||||||
|
VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER = 0x00000004u,
|
||||||
};
|
};
|
||||||
|
|
||||||
struct vkd3d_shader_interface_info
|
struct vkd3d_shader_interface_info
|
||||||
|
@ -205,6 +206,8 @@ struct vkd3d_shader_interface_info
|
||||||
|
|
||||||
/* Ignored unless VKD3D_SHADER_INTERFACE_PUSH_CONSTANTS_AS_UNIFORM_BUFFER is set */
|
/* Ignored unless VKD3D_SHADER_INTERFACE_PUSH_CONSTANTS_AS_UNIFORM_BUFFER is set */
|
||||||
const struct vkd3d_shader_descriptor_binding *push_constant_ubo_binding;
|
const struct vkd3d_shader_descriptor_binding *push_constant_ubo_binding;
|
||||||
|
/* Ignored unless VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER is set */
|
||||||
|
const struct vkd3d_shader_descriptor_binding *offset_buffer_binding;
|
||||||
};
|
};
|
||||||
|
|
||||||
struct vkd3d_shader_transform_feedback_element
|
struct vkd3d_shader_transform_feedback_element
|
||||||
|
|
|
@ -127,21 +127,19 @@ static dxil_spv_bool dxil_srv_remap(void *userdata, const dxil_spv_d3d_binding *
|
||||||
{
|
{
|
||||||
const struct vkd3d_shader_interface_info *shader_interface_info = userdata;
|
const struct vkd3d_shader_interface_info *shader_interface_info = userdata;
|
||||||
unsigned int resource_flags, resource_flags_ssbo;
|
unsigned int resource_flags, resource_flags_ssbo;
|
||||||
bool use_ssbo;
|
|
||||||
|
|
||||||
resource_flags_ssbo = dxil_resource_flags_from_kind(d3d_binding->kind, true);
|
resource_flags_ssbo = dxil_resource_flags_from_kind(d3d_binding->kind, true);
|
||||||
resource_flags = dxil_resource_flags_from_kind(d3d_binding->kind, false);
|
resource_flags = dxil_resource_flags_from_kind(d3d_binding->kind, false);
|
||||||
|
bool use_ssbo = resource_flags_ssbo != resource_flags;
|
||||||
use_ssbo = resource_flags_ssbo != resource_flags;
|
|
||||||
|
|
||||||
if (use_ssbo && dxil_remap(shader_interface_info, VKD3D_SHADER_DESCRIPTOR_TYPE_SRV,
|
if (use_ssbo && dxil_remap(shader_interface_info, VKD3D_SHADER_DESCRIPTOR_TYPE_SRV,
|
||||||
d3d_binding, &vk_binding->buffer_binding, resource_flags_ssbo))
|
d3d_binding, &vk_binding->buffer_binding, resource_flags_ssbo))
|
||||||
{
|
{
|
||||||
vk_binding->buffer_binding.descriptor_type = DXIL_SPV_VULKAN_DESCRIPTOR_TYPE_SSBO;
|
vk_binding->buffer_binding.descriptor_type = DXIL_SPV_VULKAN_DESCRIPTOR_TYPE_SSBO;
|
||||||
if (d3d_binding->alignment < shader_interface_info->min_ssbo_alignment)
|
if (shader_interface_info->flags & VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER)
|
||||||
{
|
{
|
||||||
FIXME("Shader declares resource with alignment of %u bytes, but implementation only supports %u.\n",
|
vk_binding->offset_binding.set = shader_interface_info->offset_buffer_binding->set;
|
||||||
d3d_binding->alignment, shader_interface_info->min_ssbo_alignment);
|
vk_binding->offset_binding.binding = shader_interface_info->offset_buffer_binding->binding;
|
||||||
}
|
}
|
||||||
return DXIL_SPV_TRUE;
|
return DXIL_SPV_TRUE;
|
||||||
}
|
}
|
||||||
|
@ -226,15 +224,10 @@ static dxil_spv_bool dxil_uav_remap(void *userdata, const dxil_spv_uav_d3d_bindi
|
||||||
{
|
{
|
||||||
const struct vkd3d_shader_interface_info *shader_interface_info = userdata;
|
const struct vkd3d_shader_interface_info *shader_interface_info = userdata;
|
||||||
unsigned int resource_flags, resource_flags_ssbo;
|
unsigned int resource_flags, resource_flags_ssbo;
|
||||||
bool use_ssbo;
|
|
||||||
|
|
||||||
resource_flags_ssbo = dxil_resource_flags_from_kind(d3d_binding->d3d_binding.kind, true);
|
resource_flags_ssbo = dxil_resource_flags_from_kind(d3d_binding->d3d_binding.kind, true);
|
||||||
resource_flags = dxil_resource_flags_from_kind(d3d_binding->d3d_binding.kind, false);
|
resource_flags = dxil_resource_flags_from_kind(d3d_binding->d3d_binding.kind, false);
|
||||||
|
bool use_ssbo = resource_flags != resource_flags_ssbo;
|
||||||
if (resource_flags != resource_flags_ssbo)
|
|
||||||
use_ssbo = d3d_binding->d3d_binding.alignment >= shader_interface_info->min_ssbo_alignment;
|
|
||||||
else
|
|
||||||
use_ssbo = false;
|
|
||||||
|
|
||||||
if (use_ssbo)
|
if (use_ssbo)
|
||||||
{
|
{
|
||||||
|
@ -242,6 +235,11 @@ static dxil_spv_bool dxil_uav_remap(void *userdata, const dxil_spv_uav_d3d_bindi
|
||||||
&vk_binding->buffer_binding, resource_flags_ssbo))
|
&vk_binding->buffer_binding, resource_flags_ssbo))
|
||||||
{
|
{
|
||||||
vk_binding->buffer_binding.descriptor_type = DXIL_SPV_VULKAN_DESCRIPTOR_TYPE_SSBO;
|
vk_binding->buffer_binding.descriptor_type = DXIL_SPV_VULKAN_DESCRIPTOR_TYPE_SSBO;
|
||||||
|
if (shader_interface_info->flags & VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER)
|
||||||
|
{
|
||||||
|
vk_binding->offset_binding.set = shader_interface_info->offset_buffer_binding->set;
|
||||||
|
vk_binding->offset_binding.binding = shader_interface_info->offset_buffer_binding->binding;
|
||||||
|
}
|
||||||
}
|
}
|
||||||
else if (!dxil_remap(shader_interface_info, VKD3D_SHADER_DESCRIPTOR_TYPE_UAV, &d3d_binding->d3d_binding,
|
else if (!dxil_remap(shader_interface_info, VKD3D_SHADER_DESCRIPTOR_TYPE_UAV, &d3d_binding->d3d_binding,
|
||||||
&vk_binding->buffer_binding, resource_flags))
|
&vk_binding->buffer_binding, resource_flags))
|
||||||
|
|
|
@ -2195,6 +2195,8 @@ struct vkd3d_dxbc_compiler
|
||||||
struct vkd3d_shader_global_binding *global_bindings;
|
struct vkd3d_shader_global_binding *global_bindings;
|
||||||
size_t global_bindings_size;
|
size_t global_bindings_size;
|
||||||
size_t global_binding_count;
|
size_t global_binding_count;
|
||||||
|
|
||||||
|
uint32_t offset_buffer_var_id;
|
||||||
};
|
};
|
||||||
|
|
||||||
static bool shader_is_sm_5_1(const struct vkd3d_dxbc_compiler *compiler)
|
static bool shader_is_sm_5_1(const struct vkd3d_dxbc_compiler *compiler)
|
||||||
|
@ -5446,6 +5448,33 @@ static void vkd3d_dxbc_compiler_emit_dcl_indexable_temp(struct vkd3d_dxbc_compil
|
||||||
vkd3d_dxbc_compiler_put_symbol(compiler, ®_symbol);
|
vkd3d_dxbc_compiler_put_symbol(compiler, ®_symbol);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
static void vkd3d_dxbc_compiler_emit_offset_buffer(struct vkd3d_dxbc_compiler *compiler)
|
||||||
|
{
|
||||||
|
const struct vkd3d_shader_interface_info *shader_interface = &compiler->shader_interface;
|
||||||
|
struct vkd3d_spirv_builder *builder = &compiler->spirv_builder;
|
||||||
|
uint32_t array_id, struct_id, pointer_id, var_id;
|
||||||
|
|
||||||
|
if (!(shader_interface->flags & VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER))
|
||||||
|
return;
|
||||||
|
|
||||||
|
array_id = vkd3d_spirv_build_op_type_runtime_array(builder,
|
||||||
|
vkd3d_spirv_get_type_id(builder, VKD3D_TYPE_UINT, 2));
|
||||||
|
vkd3d_spirv_build_op_decorate1(builder, array_id, SpvDecorationArrayStride, 8);
|
||||||
|
|
||||||
|
struct_id = vkd3d_spirv_build_op_type_struct(builder, &array_id, 1);
|
||||||
|
vkd3d_spirv_build_op_decorate(builder, struct_id, SpvDecorationBufferBlock, NULL, 0);
|
||||||
|
vkd3d_spirv_build_op_member_decorate1(builder, struct_id, 0, SpvDecorationOffset, 0);
|
||||||
|
vkd3d_spirv_build_op_name(builder, struct_id, "offset_buf");
|
||||||
|
|
||||||
|
pointer_id = vkd3d_spirv_build_op_type_pointer(builder, SpvStorageClassUniform, struct_id);
|
||||||
|
var_id = vkd3d_spirv_build_op_variable(builder, &builder->global_stream, pointer_id, SpvStorageClassUniform, 0);
|
||||||
|
vkd3d_spirv_build_op_decorate1(builder, var_id, SpvDecorationDescriptorSet, shader_interface->offset_buffer_binding->set);
|
||||||
|
vkd3d_spirv_build_op_decorate1(builder, var_id, SpvDecorationBinding, shader_interface->offset_buffer_binding->binding);
|
||||||
|
vkd3d_spirv_build_op_decorate(builder, var_id, SpvDecorationNonWritable, NULL, 0);
|
||||||
|
|
||||||
|
compiler->offset_buffer_var_id = var_id;
|
||||||
|
}
|
||||||
|
|
||||||
static void vkd3d_dxbc_compiler_emit_push_constant_buffers(struct vkd3d_dxbc_compiler *compiler)
|
static void vkd3d_dxbc_compiler_emit_push_constant_buffers(struct vkd3d_dxbc_compiler *compiler)
|
||||||
{
|
{
|
||||||
uint32_t uint_id, float_id, struct_id, pointer_type_id, var_id;
|
uint32_t uint_id, float_id, struct_id, pointer_type_id, var_id;
|
||||||
|
@ -8416,6 +8445,77 @@ static uint32_t vkd3d_dxbc_compiler_emit_raw_structured_addressing(
|
||||||
return offset_id;
|
return offset_id;
|
||||||
}
|
}
|
||||||
|
|
||||||
|
static uint32_t vkd3d_dxbc_compiler_get_ssbo_bounds(struct vkd3d_dxbc_compiler *compiler,
|
||||||
|
const struct vkd3d_shader_register *reg, const struct vkd3d_shader_resource_binding *binding)
|
||||||
|
{
|
||||||
|
struct vkd3d_spirv_builder *builder = &compiler->spirv_builder;
|
||||||
|
uint32_t bounds_id, vec2_ptr_id, vec2_type_id;
|
||||||
|
uint32_t indices[2];
|
||||||
|
|
||||||
|
vec2_type_id = vkd3d_spirv_get_type_id(builder, VKD3D_TYPE_UINT, 2);
|
||||||
|
vec2_ptr_id = vkd3d_spirv_get_op_type_pointer(builder, SpvStorageClassUniform, vec2_type_id);
|
||||||
|
|
||||||
|
indices[0] = vkd3d_dxbc_compiler_get_constant_uint(compiler, 0);
|
||||||
|
indices[1] = vkd3d_dxbc_compiler_get_resource_index(compiler, reg, binding);
|
||||||
|
|
||||||
|
/* returns (offset, length) in bytes */
|
||||||
|
bounds_id = vkd3d_spirv_build_op_load(builder, vec2_type_id,
|
||||||
|
vkd3d_spirv_build_op_access_chain(builder, vec2_ptr_id,
|
||||||
|
compiler->offset_buffer_var_id, indices, ARRAY_SIZE(indices)),
|
||||||
|
SpvMemoryAccessMaskNone);
|
||||||
|
|
||||||
|
return bounds_id;
|
||||||
|
}
|
||||||
|
|
||||||
|
static uint32_t vkd3d_dxbc_compiler_adjust_ssbo_offset(struct vkd3d_dxbc_compiler *compiler,
|
||||||
|
const struct vkd3d_shader_register *reg, uint32_t coordinate_id)
|
||||||
|
{
|
||||||
|
struct vkd3d_spirv_builder *builder = &compiler->spirv_builder;
|
||||||
|
uint32_t shift_id, offset_id, length_id, bounds_id, cond_id;
|
||||||
|
uint32_t uint_type_id, bool_type_id;
|
||||||
|
const struct vkd3d_symbol *symbol;
|
||||||
|
unsigned int alignment;
|
||||||
|
|
||||||
|
if (!(compiler->shader_interface.flags & VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER))
|
||||||
|
return coordinate_id;
|
||||||
|
|
||||||
|
symbol = vkd3d_dxbc_compiler_find_resource(compiler, reg);
|
||||||
|
|
||||||
|
if (symbol->info.resource.raw)
|
||||||
|
alignment = 16;
|
||||||
|
else
|
||||||
|
alignment = 4 * (symbol->info.resource.structure_stride & -symbol->info.resource.structure_stride);
|
||||||
|
|
||||||
|
/* Assume that offset is 0 and size matches the descriptor size */
|
||||||
|
if (alignment >= compiler->shader_interface.min_ssbo_alignment)
|
||||||
|
return coordinate_id;
|
||||||
|
|
||||||
|
bool_type_id = vkd3d_spirv_get_type_id(builder, VKD3D_TYPE_BOOL, 1);
|
||||||
|
uint_type_id = vkd3d_spirv_get_type_id(builder, VKD3D_TYPE_UINT, 1);
|
||||||
|
|
||||||
|
bounds_id = vkd3d_dxbc_compiler_get_ssbo_bounds(compiler, reg, symbol->info.resource.resource_binding);
|
||||||
|
|
||||||
|
shift_id = vkd3d_dxbc_compiler_get_constant_uint(compiler, 2);
|
||||||
|
offset_id = vkd3d_spirv_build_op_shift_right_logical(builder, uint_type_id,
|
||||||
|
vkd3d_spirv_build_op_composite_extract1(builder, uint_type_id, bounds_id, 0), shift_id);
|
||||||
|
length_id = vkd3d_spirv_build_op_shift_right_logical(builder, uint_type_id,
|
||||||
|
vkd3d_spirv_build_op_composite_extract1(builder, uint_type_id, bounds_id, 1), shift_id);
|
||||||
|
|
||||||
|
/* cond = offset < length */
|
||||||
|
cond_id = vkd3d_spirv_build_op_uless_than(builder, bool_type_id, coordinate_id, length_id);
|
||||||
|
|
||||||
|
/* In case of out-of-bounds access, set offset to a number that we
|
||||||
|
* expect to be out-of-bounds of the actual Vulkan resource as well.
|
||||||
|
* 0x3ffffffc is the largest offset value we can safely use without
|
||||||
|
* overflowing 32-bit address space, since this is a DWORD offset
|
||||||
|
* and we may access a total of 16 bytes starting at that offset. */
|
||||||
|
coordinate_id = vkd3d_spirv_build_op_select(builder, uint_type_id, cond_id,
|
||||||
|
vkd3d_spirv_build_op_iadd(builder, uint_type_id, coordinate_id, offset_id),
|
||||||
|
vkd3d_dxbc_compiler_get_constant_uint(compiler, 0x3ffffffc));
|
||||||
|
|
||||||
|
return coordinate_id;
|
||||||
|
}
|
||||||
|
|
||||||
static void vkd3d_dxbc_compiler_emit_ld_raw_structured_srv_uav(struct vkd3d_dxbc_compiler *compiler,
|
static void vkd3d_dxbc_compiler_emit_ld_raw_structured_srv_uav(struct vkd3d_dxbc_compiler *compiler,
|
||||||
const struct vkd3d_shader_instruction *instruction)
|
const struct vkd3d_shader_instruction *instruction)
|
||||||
{
|
{
|
||||||
|
@ -8462,6 +8562,9 @@ static void vkd3d_dxbc_compiler_emit_ld_raw_structured_srv_uav(struct vkd3d_dxbc
|
||||||
base_coordinate_id = vkd3d_dxbc_compiler_emit_raw_structured_addressing(compiler,
|
base_coordinate_id = vkd3d_dxbc_compiler_emit_raw_structured_addressing(compiler,
|
||||||
type_id, image.structure_stride, &src[0], VKD3DSP_WRITEMASK_0, &src[1], VKD3DSP_WRITEMASK_0);
|
type_id, image.structure_stride, &src[0], VKD3DSP_WRITEMASK_0, &src[1], VKD3DSP_WRITEMASK_0);
|
||||||
|
|
||||||
|
if (image.ssbo)
|
||||||
|
base_coordinate_id = vkd3d_dxbc_compiler_adjust_ssbo_offset(compiler, &resource->reg, base_coordinate_id);
|
||||||
|
|
||||||
texel_type_id = vkd3d_spirv_get_type_id(builder, image.sampled_type, VKD3D_VEC4_SIZE);
|
texel_type_id = vkd3d_spirv_get_type_id(builder, image.sampled_type, VKD3D_VEC4_SIZE);
|
||||||
result_type_id = is_sparse_op ? vkd3d_spirv_get_sparse_result_type(builder, texel_type_id) : texel_type_id;
|
result_type_id = is_sparse_op ? vkd3d_spirv_get_sparse_result_type(builder, texel_type_id) : texel_type_id;
|
||||||
assert(dst->write_mask & VKD3DSP_WRITEMASK_ALL);
|
assert(dst->write_mask & VKD3DSP_WRITEMASK_ALL);
|
||||||
|
@ -8594,6 +8697,9 @@ static void vkd3d_dxbc_compiler_emit_store_uav_raw_structured(struct vkd3d_dxbc_
|
||||||
base_coordinate_id = vkd3d_dxbc_compiler_emit_raw_structured_addressing(compiler,
|
base_coordinate_id = vkd3d_dxbc_compiler_emit_raw_structured_addressing(compiler,
|
||||||
type_id, image.structure_stride, &src[0], VKD3DSP_WRITEMASK_0, &src[1], VKD3DSP_WRITEMASK_0);
|
type_id, image.structure_stride, &src[0], VKD3DSP_WRITEMASK_0, &src[1], VKD3DSP_WRITEMASK_0);
|
||||||
|
|
||||||
|
if (image.ssbo)
|
||||||
|
base_coordinate_id = vkd3d_dxbc_compiler_adjust_ssbo_offset(compiler, &dst->reg, base_coordinate_id);
|
||||||
|
|
||||||
texel = &src[instruction->src_count - 1];
|
texel = &src[instruction->src_count - 1];
|
||||||
assert(texel->reg.data_type == VKD3D_DATA_UINT);
|
assert(texel->reg.data_type == VKD3D_DATA_UINT);
|
||||||
val_id = vkd3d_dxbc_compiler_emit_load_src(compiler, texel, dst->write_mask);
|
val_id = vkd3d_dxbc_compiler_emit_load_src(compiler, texel, dst->write_mask);
|
||||||
|
@ -8936,6 +9042,9 @@ static void vkd3d_dxbc_compiler_emit_atomic_instruction(struct vkd3d_dxbc_compil
|
||||||
coordinate_id = vkd3d_dxbc_compiler_emit_raw_structured_addressing(compiler,
|
coordinate_id = vkd3d_dxbc_compiler_emit_raw_structured_addressing(compiler,
|
||||||
type_id, structure_stride, &src[0], VKD3DSP_WRITEMASK_0,
|
type_id, structure_stride, &src[0], VKD3DSP_WRITEMASK_0,
|
||||||
&src[0], VKD3DSP_WRITEMASK_1);
|
&src[0], VKD3DSP_WRITEMASK_1);
|
||||||
|
|
||||||
|
if (resource->reg.type != VKD3DSPR_GROUPSHAREDMEM && image.ssbo)
|
||||||
|
coordinate_id = vkd3d_dxbc_compiler_adjust_ssbo_offset(compiler, &resource->reg, coordinate_id);
|
||||||
}
|
}
|
||||||
else
|
else
|
||||||
{
|
{
|
||||||
|
@ -8999,7 +9108,7 @@ static void vkd3d_dxbc_compiler_emit_bufinfo(struct vkd3d_dxbc_compiler *compile
|
||||||
struct vkd3d_spirv_builder *builder = &compiler->spirv_builder;
|
struct vkd3d_spirv_builder *builder = &compiler->spirv_builder;
|
||||||
const struct vkd3d_shader_dst_param *dst = instruction->dst;
|
const struct vkd3d_shader_dst_param *dst = instruction->dst;
|
||||||
const struct vkd3d_shader_src_param *src = instruction->src;
|
const struct vkd3d_shader_src_param *src = instruction->src;
|
||||||
uint32_t type_id, val_id, stride_id;
|
uint32_t type_id, val_id, stride_id, bounds_id;
|
||||||
struct vkd3d_shader_image image;
|
struct vkd3d_shader_image image;
|
||||||
uint32_t constituents[2];
|
uint32_t constituents[2];
|
||||||
unsigned int write_mask;
|
unsigned int write_mask;
|
||||||
|
@ -9012,10 +9121,22 @@ static void vkd3d_dxbc_compiler_emit_bufinfo(struct vkd3d_dxbc_compiler *compile
|
||||||
|
|
||||||
if (image.ssbo)
|
if (image.ssbo)
|
||||||
{
|
{
|
||||||
if (src->reg.modifier == VKD3DSPRM_NONUNIFORM)
|
if (compiler->shader_interface.flags & VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER)
|
||||||
vkd3d_dxbc_compiler_decorate_nonuniform(compiler, image.id);
|
{
|
||||||
|
const struct vkd3d_symbol *symbol = vkd3d_dxbc_compiler_find_resource(compiler, &src->reg);
|
||||||
|
bounds_id = vkd3d_dxbc_compiler_get_ssbo_bounds(compiler, &src->reg, symbol->info.resource.resource_binding);
|
||||||
|
|
||||||
val_id = vkd3d_spirv_build_op_array_length(builder, type_id, image.id, 0);
|
val_id = vkd3d_spirv_build_op_shift_right_logical(builder, type_id,
|
||||||
|
vkd3d_spirv_build_op_composite_extract1(builder, type_id, bounds_id, 1),
|
||||||
|
vkd3d_dxbc_compiler_get_constant_uint(compiler, 2));
|
||||||
|
}
|
||||||
|
else
|
||||||
|
{
|
||||||
|
if (src->reg.modifier == VKD3DSPRM_NONUNIFORM)
|
||||||
|
vkd3d_dxbc_compiler_decorate_nonuniform(compiler, image.id);
|
||||||
|
|
||||||
|
val_id = vkd3d_spirv_build_op_array_length(builder, type_id, image.id, 0);
|
||||||
|
}
|
||||||
}
|
}
|
||||||
else
|
else
|
||||||
val_id = vkd3d_spirv_build_op_image_query_size(builder, type_id, image.image_id);
|
val_id = vkd3d_spirv_build_op_image_query_size(builder, type_id, image.image_id);
|
||||||
|
@ -9457,6 +9578,7 @@ static void vkd3d_dxbc_compiler_emit_double_conversion(struct vkd3d_dxbc_compile
|
||||||
/* This function is called after declarations are processed. */
|
/* This function is called after declarations are processed. */
|
||||||
static void vkd3d_dxbc_compiler_emit_main_prolog(struct vkd3d_dxbc_compiler *compiler)
|
static void vkd3d_dxbc_compiler_emit_main_prolog(struct vkd3d_dxbc_compiler *compiler)
|
||||||
{
|
{
|
||||||
|
vkd3d_dxbc_compiler_emit_offset_buffer(compiler);
|
||||||
vkd3d_dxbc_compiler_emit_push_constant_buffers(compiler);
|
vkd3d_dxbc_compiler_emit_push_constant_buffers(compiler);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
|
|
@ -5164,7 +5164,7 @@ static void d3d12_command_list_set_root_descriptor(struct d3d12_command_list *li
|
||||||
bool null_descriptors, ssbo;
|
bool null_descriptors, ssbo;
|
||||||
VkDeviceSize max_range;
|
VkDeviceSize max_range;
|
||||||
|
|
||||||
ssbo = d3d12_device_use_ssbo_raw_buffer(list->device);
|
ssbo = d3d12_device_use_ssbo_root_descriptors(list->device);
|
||||||
root_parameter = root_signature_get_root_descriptor(root_signature, index);
|
root_parameter = root_signature_get_root_descriptor(root_signature, index);
|
||||||
descriptor = &bindings->root_descriptors[root_parameter->descriptor.packed_descriptor];
|
descriptor = &bindings->root_descriptors[root_parameter->descriptor.packed_descriptor];
|
||||||
null_descriptors = list->device->device_info.robustness2_features.nullDescriptor;
|
null_descriptors = list->device->device_info.robustness2_features.nullDescriptor;
|
||||||
|
@ -5711,7 +5711,7 @@ struct vkd3d_clear_uav_info
|
||||||
} u;
|
} u;
|
||||||
};
|
};
|
||||||
|
|
||||||
static void d3d12_command_list_clear_uav(struct d3d12_command_list *list,
|
static void d3d12_command_list_clear_uav(struct d3d12_command_list *list, const struct d3d12_desc *desc,
|
||||||
struct d3d12_resource *resource, const struct vkd3d_clear_uav_info *args,
|
struct d3d12_resource *resource, const struct vkd3d_clear_uav_info *args,
|
||||||
const VkClearColorValue *clear_color, UINT rect_count, const D3D12_RECT *rects)
|
const VkClearColorValue *clear_color, UINT rect_count, const D3D12_RECT *rects)
|
||||||
{
|
{
|
||||||
|
@ -5724,6 +5724,7 @@ static void d3d12_command_list_clear_uav(struct d3d12_command_list *list,
|
||||||
D3D12_RECT full_rect, curr_rect;
|
D3D12_RECT full_rect, curr_rect;
|
||||||
VkWriteDescriptorSet write_set;
|
VkWriteDescriptorSet write_set;
|
||||||
VkExtent3D workgroup_size;
|
VkExtent3D workgroup_size;
|
||||||
|
uint32_t extra_offset;
|
||||||
|
|
||||||
d3d12_command_list_track_resource_usage(list, resource);
|
d3d12_command_list_track_resource_usage(list, resource);
|
||||||
d3d12_command_list_end_current_render_pass(list, false);
|
d3d12_command_list_end_current_render_pass(list, false);
|
||||||
|
@ -5802,6 +5803,7 @@ static void d3d12_command_list_clear_uav(struct d3d12_command_list *list,
|
||||||
full_rect.right = d3d12_resource_desc_get_width(&resource->desc, miplevel_idx);
|
full_rect.right = d3d12_resource_desc_get_width(&resource->desc, miplevel_idx);
|
||||||
full_rect.top = 0;
|
full_rect.top = 0;
|
||||||
full_rect.bottom = d3d12_resource_desc_get_height(&resource->desc, miplevel_idx);
|
full_rect.bottom = d3d12_resource_desc_get_height(&resource->desc, miplevel_idx);
|
||||||
|
extra_offset = 0;
|
||||||
|
|
||||||
if (d3d12_resource_is_buffer(resource))
|
if (d3d12_resource_is_buffer(resource))
|
||||||
{
|
{
|
||||||
|
@ -5812,6 +5814,12 @@ static void d3d12_command_list_clear_uav(struct d3d12_command_list *list,
|
||||||
: sizeof(uint32_t); /* structured buffer */
|
: sizeof(uint32_t); /* structured buffer */
|
||||||
full_rect.right = args->u.view->info.buffer.size / byte_count;
|
full_rect.right = args->u.view->info.buffer.size / byte_count;
|
||||||
}
|
}
|
||||||
|
else if (list->device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
{
|
||||||
|
const struct vkd3d_bound_ssbo_range *ranges = desc->heap->ssbo_ranges.host_ptr;
|
||||||
|
extra_offset = ranges[desc->heap_offset].offset / sizeof(uint32_t);
|
||||||
|
full_rect.right = ranges[desc->heap_offset].length / sizeof(uint32_t);
|
||||||
|
}
|
||||||
else
|
else
|
||||||
full_rect.right = args->u.buffer.range / sizeof(uint32_t);
|
full_rect.right = args->u.buffer.range / sizeof(uint32_t);
|
||||||
}
|
}
|
||||||
|
@ -5840,7 +5848,7 @@ static void d3d12_command_list_clear_uav(struct d3d12_command_list *list,
|
||||||
continue;
|
continue;
|
||||||
}
|
}
|
||||||
|
|
||||||
clear_args.offset.x = curr_rect.left;
|
clear_args.offset.x = curr_rect.left + extra_offset;
|
||||||
clear_args.offset.y = curr_rect.top;
|
clear_args.offset.y = curr_rect.top;
|
||||||
clear_args.extent.width = curr_rect.right - curr_rect.left;
|
clear_args.extent.width = curr_rect.right - curr_rect.left;
|
||||||
clear_args.extent.height = curr_rect.bottom - curr_rect.top;
|
clear_args.extent.height = curr_rect.bottom - curr_rect.top;
|
||||||
|
@ -5956,7 +5964,7 @@ static void STDMETHODCALLTYPE d3d12_command_list_ClearUnorderedAccessViewUint(d3
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
d3d12_command_list_clear_uav(list, resource_impl, &args, &color, rect_count, rects);
|
d3d12_command_list_clear_uav(list, desc, resource_impl, &args, &color, rect_count, rects);
|
||||||
}
|
}
|
||||||
|
|
||||||
static void STDMETHODCALLTYPE d3d12_command_list_ClearUnorderedAccessViewFloat(d3d12_command_list_iface *iface,
|
static void STDMETHODCALLTYPE d3d12_command_list_ClearUnorderedAccessViewFloat(d3d12_command_list_iface *iface,
|
||||||
|
@ -5977,7 +5985,7 @@ static void STDMETHODCALLTYPE d3d12_command_list_ClearUnorderedAccessViewFloat(d
|
||||||
resource_impl = unsafe_impl_from_ID3D12Resource(resource);
|
resource_impl = unsafe_impl_from_ID3D12Resource(resource);
|
||||||
|
|
||||||
vkd3d_clear_uav_info_from_desc(&args, desc);
|
vkd3d_clear_uav_info_from_desc(&args, desc);
|
||||||
d3d12_command_list_clear_uav(list, resource_impl, &args, &color, rect_count, rects);
|
d3d12_command_list_clear_uav(list, desc, resource_impl, &args, &color, rect_count, rects);
|
||||||
}
|
}
|
||||||
|
|
||||||
static void STDMETHODCALLTYPE d3d12_command_list_DiscardResource(d3d12_command_list_iface *iface,
|
static void STDMETHODCALLTYPE d3d12_command_list_DiscardResource(d3d12_command_list_iface *iface,
|
||||||
|
|
|
@ -3478,9 +3478,10 @@ void d3d12_desc_copy(struct d3d12_desc *dst, struct d3d12_desc *src,
|
||||||
|
|
||||||
if (metadata.flags & VKD3D_DESCRIPTOR_FLAG_UAV_COUNTER)
|
if (metadata.flags & VKD3D_DESCRIPTOR_FLAG_UAV_COUNTER)
|
||||||
{
|
{
|
||||||
if (dst->heap->uav_counters.data)
|
if (dst->heap->uav_counters.host_ptr)
|
||||||
{
|
{
|
||||||
dst->heap->uav_counters.data[dst->heap_offset] = src->counter_address;
|
VkDeviceAddress *counter_addresses = dst->heap->uav_counters.host_ptr;
|
||||||
|
counter_addresses[dst->heap_offset] = src->counter_address;
|
||||||
dst->counter_address = src->counter_address;
|
dst->counter_address = src->counter_address;
|
||||||
}
|
}
|
||||||
else
|
else
|
||||||
|
@ -3501,6 +3502,13 @@ void d3d12_desc_copy(struct d3d12_desc *dst, struct d3d12_desc *src,
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
|
if (metadata.flags & VKD3D_DESCRIPTOR_FLAG_SSBO_OFFSET)
|
||||||
|
{
|
||||||
|
const struct vkd3d_bound_ssbo_range *src_ssbo_ranges = src->heap->ssbo_ranges.host_ptr;
|
||||||
|
struct vkd3d_bound_ssbo_range *dst_ssbo_ranges = dst->heap->ssbo_ranges.host_ptr;
|
||||||
|
dst_ssbo_ranges[dst->heap_offset] = src_ssbo_ranges[src->heap_offset];
|
||||||
|
}
|
||||||
|
|
||||||
if (copy_count)
|
if (copy_count)
|
||||||
VK_CALL(vkUpdateDescriptorSets(device->vk_device, 0, NULL, copy_count, vk_copies));
|
VK_CALL(vkUpdateDescriptorSets(device->vk_device, 0, NULL, copy_count, vk_copies));
|
||||||
}
|
}
|
||||||
|
@ -3947,6 +3955,43 @@ static bool vkd3d_buffer_srv_use_raw_ssbo(struct d3d12_device *device,
|
||||||
((desc->Format == DXGI_FORMAT_UNKNOWN && desc->Buffer.StructureByteStride) || raw);
|
((desc->Format == DXGI_FORMAT_UNKNOWN && desc->Buffer.StructureByteStride) || raw);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
static void vkd3d_buffer_view_get_bound_range(struct d3d12_desc *descriptor,
|
||||||
|
struct d3d12_device *device, struct d3d12_resource *resource,
|
||||||
|
VkDeviceSize offset, VkDeviceSize range, VkDescriptorBufferInfo *vk_buffer)
|
||||||
|
{
|
||||||
|
struct vkd3d_bound_ssbo_range ssbo_range;
|
||||||
|
|
||||||
|
if (resource)
|
||||||
|
{
|
||||||
|
VkDeviceSize alignment = d3d12_device_get_ssbo_alignment(device);
|
||||||
|
VkDeviceSize aligned_begin = offset & ~(alignment - 1);
|
||||||
|
VkDeviceSize aligned_end = min((offset + range + alignment - 1) & ~(alignment - 1), resource->desc.Width);
|
||||||
|
|
||||||
|
/* heap_offset is guaranteed to have 64KiB alignment */
|
||||||
|
vk_buffer->buffer = resource->vk_buffer;
|
||||||
|
vk_buffer->offset = resource->heap_offset + aligned_begin;
|
||||||
|
vk_buffer->range = aligned_end - aligned_begin;
|
||||||
|
|
||||||
|
ssbo_range.offset = offset - aligned_begin;
|
||||||
|
ssbo_range.length = range;
|
||||||
|
}
|
||||||
|
else
|
||||||
|
{
|
||||||
|
vk_buffer->buffer = VK_NULL_HANDLE;
|
||||||
|
vk_buffer->offset = 0;
|
||||||
|
vk_buffer->range = 0;
|
||||||
|
|
||||||
|
ssbo_range.offset = 0;
|
||||||
|
ssbo_range.length = 0;
|
||||||
|
}
|
||||||
|
|
||||||
|
if (device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
{
|
||||||
|
struct vkd3d_bound_ssbo_range *ssbo_ranges = descriptor->heap->ssbo_ranges.host_ptr;
|
||||||
|
ssbo_ranges[descriptor->heap_offset] = ssbo_range;
|
||||||
|
}
|
||||||
|
}
|
||||||
|
|
||||||
static void vkd3d_create_buffer_srv(struct d3d12_desc *descriptor,
|
static void vkd3d_create_buffer_srv(struct d3d12_desc *descriptor,
|
||||||
struct d3d12_device *device, struct d3d12_resource *resource,
|
struct d3d12_device *device, struct d3d12_resource *resource,
|
||||||
const D3D12_SHADER_RESOURCE_VIEW_DESC *desc)
|
const D3D12_SHADER_RESOURCE_VIEW_DESC *desc)
|
||||||
|
@ -3973,26 +4018,12 @@ static void vkd3d_create_buffer_srv(struct d3d12_desc *descriptor,
|
||||||
|
|
||||||
if (vkd3d_buffer_srv_use_raw_ssbo(device, desc))
|
if (vkd3d_buffer_srv_use_raw_ssbo(device, desc))
|
||||||
{
|
{
|
||||||
if (resource)
|
VkDeviceSize stride = desc->Format == DXGI_FORMAT_UNKNOWN
|
||||||
{
|
? desc->Buffer.StructureByteStride : sizeof(uint32_t);
|
||||||
uint32_t stride = desc->Format == DXGI_FORMAT_UNKNOWN
|
|
||||||
? desc->Buffer.StructureByteStride : sizeof(uint32_t);
|
|
||||||
descriptor_info.buffer.buffer = resource->vk_buffer;
|
|
||||||
descriptor_info.buffer.offset = desc->Buffer.FirstElement * stride + resource->heap_offset;
|
|
||||||
descriptor_info.buffer.range = desc->Buffer.NumElements * stride;
|
|
||||||
|
|
||||||
if (descriptor_info.buffer.offset & (d3d12_device_get_ssbo_alignment(device) - 1))
|
vkd3d_buffer_view_get_bound_range(descriptor, device, resource,
|
||||||
{
|
desc->Buffer.FirstElement * stride, desc->Buffer.NumElements * stride,
|
||||||
FIXME("Emitting SSBO at offset #%"PRIx64", but needs alignment of %"PRIu64" bytes.\n",
|
&descriptor_info.buffer);
|
||||||
descriptor_info.buffer.offset, d3d12_device_get_ssbo_alignment(device));
|
|
||||||
}
|
|
||||||
}
|
|
||||||
else
|
|
||||||
{
|
|
||||||
descriptor_info.buffer.buffer = VK_NULL_HANDLE;
|
|
||||||
descriptor_info.buffer.offset = 0;
|
|
||||||
descriptor_info.buffer.range = 0;
|
|
||||||
}
|
|
||||||
|
|
||||||
descriptor->info.buffer = descriptor_info.buffer;
|
descriptor->info.buffer = descriptor_info.buffer;
|
||||||
descriptor->metadata.cookie = resource ? resource->cookie : 0;
|
descriptor->metadata.cookie = resource ? resource->cookie : 0;
|
||||||
|
@ -4000,6 +4031,9 @@ static void vkd3d_create_buffer_srv(struct d3d12_desc *descriptor,
|
||||||
VKD3D_BINDLESS_SET_SRV | VKD3D_BINDLESS_SET_RAW_SSBO);
|
VKD3D_BINDLESS_SET_SRV | VKD3D_BINDLESS_SET_RAW_SSBO);
|
||||||
descriptor->metadata.flags = VKD3D_DESCRIPTOR_FLAG_DEFINED;
|
descriptor->metadata.flags = VKD3D_DESCRIPTOR_FLAG_DEFINED;
|
||||||
|
|
||||||
|
if (device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
descriptor->metadata.flags |= VKD3D_DESCRIPTOR_FLAG_SSBO_OFFSET;
|
||||||
|
|
||||||
vk_descriptor_type = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER;
|
vk_descriptor_type = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER;
|
||||||
}
|
}
|
||||||
else
|
else
|
||||||
|
@ -4286,33 +4320,21 @@ static void vkd3d_create_buffer_uav(struct d3d12_desc *descriptor, struct d3d12_
|
||||||
if (vkd3d_buffer_uav_use_raw_ssbo(device, desc))
|
if (vkd3d_buffer_uav_use_raw_ssbo(device, desc))
|
||||||
{
|
{
|
||||||
VkDescriptorBufferInfo *buffer_info = &descriptor_info[vk_write_count].buffer;
|
VkDescriptorBufferInfo *buffer_info = &descriptor_info[vk_write_count].buffer;
|
||||||
|
VkDeviceSize stride = desc->Format == DXGI_FORMAT_UNKNOWN
|
||||||
|
? desc->Buffer.StructureByteStride : sizeof(uint32_t);
|
||||||
|
|
||||||
if (resource)
|
vkd3d_buffer_view_get_bound_range(descriptor, device, resource,
|
||||||
{
|
desc->Buffer.FirstElement * stride, desc->Buffer.NumElements * stride,
|
||||||
uint32_t stride = desc->Format == DXGI_FORMAT_UNKNOWN
|
buffer_info);
|
||||||
? desc->Buffer.StructureByteStride : sizeof(uint32_t);
|
|
||||||
buffer_info->buffer = resource->vk_buffer;
|
|
||||||
buffer_info->offset = desc->Buffer.FirstElement * stride + resource->heap_offset;
|
|
||||||
buffer_info->range = desc->Buffer.NumElements * stride;
|
|
||||||
|
|
||||||
if (buffer_info->offset & (d3d12_device_get_ssbo_alignment(device) - 1))
|
|
||||||
{
|
|
||||||
FIXME("Emitting SSBO at offset #%"PRIx64", but needs alignment of %"PRIu64" bytes.\n",
|
|
||||||
buffer_info->offset, d3d12_device_get_ssbo_alignment(device));
|
|
||||||
}
|
|
||||||
}
|
|
||||||
else
|
|
||||||
{
|
|
||||||
buffer_info->buffer = VK_NULL_HANDLE;
|
|
||||||
buffer_info->offset = 0;
|
|
||||||
buffer_info->range = 0;
|
|
||||||
}
|
|
||||||
|
|
||||||
descriptor->info.buffer = *buffer_info;
|
descriptor->info.buffer = *buffer_info;
|
||||||
descriptor->metadata.cookie = resource ? resource->cookie : 0;
|
descriptor->metadata.cookie = resource ? resource->cookie : 0;
|
||||||
descriptor->metadata.binding = vkd3d_bindless_state_find_set(&device->bindless_state,
|
descriptor->metadata.binding = vkd3d_bindless_state_find_set(&device->bindless_state,
|
||||||
VKD3D_BINDLESS_SET_UAV | VKD3D_BINDLESS_SET_RAW_SSBO);
|
VKD3D_BINDLESS_SET_UAV | VKD3D_BINDLESS_SET_RAW_SSBO);
|
||||||
descriptor->metadata.flags = VKD3D_DESCRIPTOR_FLAG_DEFINED;
|
descriptor->metadata.flags = VKD3D_DESCRIPTOR_FLAG_DEFINED | VKD3D_DESCRIPTOR_FLAG_UAV_COUNTER;
|
||||||
|
|
||||||
|
if (device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
descriptor->metadata.flags |= VKD3D_DESCRIPTOR_FLAG_SSBO_OFFSET;
|
||||||
|
|
||||||
vk_descriptor_type = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER;
|
vk_descriptor_type = VK_DESCRIPTOR_TYPE_STORAGE_BUFFER;
|
||||||
}
|
}
|
||||||
|
@ -4382,8 +4404,9 @@ static void vkd3d_create_buffer_uav(struct d3d12_desc *descriptor, struct d3d12_
|
||||||
|
|
||||||
if (device->bindless_state.flags & VKD3D_RAW_VA_UAV_COUNTER)
|
if (device->bindless_state.flags & VKD3D_RAW_VA_UAV_COUNTER)
|
||||||
{
|
{
|
||||||
|
VkDeviceAddress *counter_addresses = descriptor->heap->uav_counters.host_ptr;
|
||||||
uint32_t descriptor_index = d3d12_desc_heap_offset(descriptor);
|
uint32_t descriptor_index = d3d12_desc_heap_offset(descriptor);
|
||||||
descriptor->heap->uav_counters.data[descriptor_index] = uav_counter_address;
|
counter_addresses[descriptor_index] = uav_counter_address;
|
||||||
descriptor->counter_address = uav_counter_address;
|
descriptor->counter_address = uav_counter_address;
|
||||||
}
|
}
|
||||||
else
|
else
|
||||||
|
@ -5400,48 +5423,98 @@ static HRESULT d3d12_descriptor_heap_create_descriptor_set(struct d3d12_descript
|
||||||
return S_OK;
|
return S_OK;
|
||||||
}
|
}
|
||||||
|
|
||||||
static HRESULT d3d12_descriptor_heap_create_uav_counter_buffer(struct d3d12_descriptor_heap *descriptor_heap,
|
static void d3d12_descriptor_heap_get_buffer_range(struct d3d12_descriptor_heap *descriptor_heap,
|
||||||
struct d3d12_descriptor_heap_uav_counters *uav_counters)
|
VkDeviceSize *offset, VkDeviceSize size, struct vkd3d_host_visible_buffer_range *range)
|
||||||
|
{
|
||||||
|
if (size)
|
||||||
|
{
|
||||||
|
range->descriptor.buffer = descriptor_heap->vk_buffer;
|
||||||
|
range->descriptor.offset = *offset;
|
||||||
|
range->descriptor.range = size;
|
||||||
|
range->host_ptr = void_ptr_offset(descriptor_heap->host_memory, *offset);
|
||||||
|
|
||||||
|
*offset += size;
|
||||||
|
}
|
||||||
|
else
|
||||||
|
{
|
||||||
|
range->descriptor.buffer = VK_NULL_HANDLE;
|
||||||
|
range->descriptor.offset = 0;
|
||||||
|
range->descriptor.range = VK_WHOLE_SIZE;
|
||||||
|
range->host_ptr = NULL;
|
||||||
|
}
|
||||||
|
}
|
||||||
|
|
||||||
|
static HRESULT d3d12_descriptor_heap_init_data_buffer(struct d3d12_descriptor_heap *descriptor_heap,
|
||||||
|
struct d3d12_device *device, const D3D12_DESCRIPTOR_HEAP_DESC *desc)
|
||||||
{
|
{
|
||||||
const struct vkd3d_vk_device_procs *vk_procs = &descriptor_heap->device->vk_procs;
|
const struct vkd3d_vk_device_procs *vk_procs = &descriptor_heap->device->vk_procs;
|
||||||
struct d3d12_device *device = descriptor_heap->device;
|
VkDeviceSize alignment = max(device->device_info.properties2.properties.limits.minStorageBufferOffsetAlignment,
|
||||||
|
device->device_info.properties2.properties.limits.nonCoherentAtomSize);
|
||||||
|
VkDeviceSize uav_counter_size = 0, offset_buffer_size = 0;
|
||||||
|
VkDeviceSize buffer_size, offset;
|
||||||
D3D12_HEAP_PROPERTIES heap_info;
|
D3D12_HEAP_PROPERTIES heap_info;
|
||||||
D3D12_RESOURCE_DESC buffer_desc;
|
D3D12_RESOURCE_DESC buffer_desc;
|
||||||
D3D12_HEAP_FLAGS heap_flags;
|
D3D12_HEAP_FLAGS heap_flags;
|
||||||
VkResult vr;
|
VkResult vr;
|
||||||
HRESULT hr;
|
HRESULT hr;
|
||||||
|
|
||||||
/* concurrently accessible storage buffer */
|
if (desc->Type == D3D12_DESCRIPTOR_HEAP_TYPE_CBV_SRV_UAV)
|
||||||
memset(&buffer_desc, 0, sizeof(buffer_desc));
|
|
||||||
buffer_desc.Dimension = D3D12_RESOURCE_DIMENSION_BUFFER;
|
|
||||||
buffer_desc.Width = descriptor_heap->desc.NumDescriptors * sizeof(VkDeviceAddress);
|
|
||||||
buffer_desc.Height = 1;
|
|
||||||
buffer_desc.DepthOrArraySize = 1;
|
|
||||||
buffer_desc.MipLevels = 1;
|
|
||||||
buffer_desc.SampleDesc.Count = 1;
|
|
||||||
buffer_desc.Layout = D3D12_TEXTURE_LAYOUT_ROW_MAJOR;
|
|
||||||
buffer_desc.Flags = D3D12_RESOURCE_FLAG_ALLOW_UNORDERED_ACCESS;
|
|
||||||
|
|
||||||
/* host-visible device memory */
|
|
||||||
memset(&heap_info, 0, sizeof(heap_info));
|
|
||||||
heap_info.Type = D3D12_HEAP_TYPE_UPLOAD;
|
|
||||||
|
|
||||||
heap_flags = D3D12_HEAP_FLAG_ALLOW_ONLY_BUFFERS;
|
|
||||||
|
|
||||||
if (FAILED(hr = vkd3d_create_buffer(device, &heap_info, heap_flags, &buffer_desc, &uav_counters->vk_buffer)))
|
|
||||||
return hr;
|
|
||||||
|
|
||||||
if (FAILED(hr = vkd3d_allocate_buffer_memory(device, uav_counters->vk_buffer, NULL,
|
|
||||||
&heap_info, heap_flags, &uav_counters->vk_memory, NULL, NULL)))
|
|
||||||
return hr;
|
|
||||||
|
|
||||||
if ((vr = VK_CALL(vkMapMemory(device->vk_device, uav_counters->vk_memory,
|
|
||||||
0, VK_WHOLE_SIZE, 0, (void **)&uav_counters->data))))
|
|
||||||
{
|
{
|
||||||
ERR("Failed to map UAV counter address buffer, vr %d.\n", vr);
|
if (device->bindless_state.flags & VKD3D_RAW_VA_UAV_COUNTER)
|
||||||
return hresult_from_vk_result(vr);
|
uav_counter_size = align(desc->NumDescriptors * sizeof(VkDeviceAddress), alignment);
|
||||||
|
|
||||||
|
if (device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
offset_buffer_size = align(desc->NumDescriptors * sizeof(struct vkd3d_bound_ssbo_range), alignment);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
buffer_size = uav_counter_size + offset_buffer_size;
|
||||||
|
|
||||||
|
if (!buffer_size)
|
||||||
|
return S_OK;
|
||||||
|
|
||||||
|
if (desc->Flags & D3D12_DESCRIPTOR_HEAP_FLAG_SHADER_VISIBLE)
|
||||||
|
{
|
||||||
|
memset(&buffer_desc, 0, sizeof(buffer_desc));
|
||||||
|
buffer_desc.Dimension = D3D12_RESOURCE_DIMENSION_BUFFER;
|
||||||
|
buffer_desc.Width = buffer_size;
|
||||||
|
buffer_desc.Height = 1;
|
||||||
|
buffer_desc.DepthOrArraySize = 1;
|
||||||
|
buffer_desc.MipLevels = 1;
|
||||||
|
buffer_desc.SampleDesc.Count = 1;
|
||||||
|
buffer_desc.Layout = D3D12_TEXTURE_LAYOUT_ROW_MAJOR;
|
||||||
|
buffer_desc.Flags = D3D12_RESOURCE_FLAG_ALLOW_UNORDERED_ACCESS;
|
||||||
|
|
||||||
|
/* host-visible device memory */
|
||||||
|
memset(&heap_info, 0, sizeof(heap_info));
|
||||||
|
heap_info.Type = D3D12_HEAP_TYPE_UPLOAD;
|
||||||
|
|
||||||
|
heap_flags = D3D12_HEAP_FLAG_ALLOW_ONLY_BUFFERS;
|
||||||
|
|
||||||
|
if (FAILED(hr = vkd3d_create_buffer(device, &heap_info, heap_flags, &buffer_desc, &descriptor_heap->vk_buffer)))
|
||||||
|
return hr;
|
||||||
|
|
||||||
|
if (FAILED(hr = vkd3d_allocate_buffer_memory(device, descriptor_heap->vk_buffer, NULL,
|
||||||
|
&heap_info, heap_flags, &descriptor_heap->vk_memory, NULL, NULL)))
|
||||||
|
return hr;
|
||||||
|
|
||||||
|
if ((vr = VK_CALL(vkMapMemory(device->vk_device, descriptor_heap->vk_memory,
|
||||||
|
0, VK_WHOLE_SIZE, 0, &descriptor_heap->host_memory))))
|
||||||
|
{
|
||||||
|
ERR("Failed to map buffer, vr %d.\n", vr);
|
||||||
|
return hresult_from_vk_result(vr);
|
||||||
|
}
|
||||||
|
}
|
||||||
|
else
|
||||||
|
{
|
||||||
|
descriptor_heap->vk_memory = VK_NULL_HANDLE;
|
||||||
|
descriptor_heap->vk_buffer = VK_NULL_HANDLE;
|
||||||
|
descriptor_heap->host_memory = vkd3d_calloc(1, buffer_size);
|
||||||
|
}
|
||||||
|
|
||||||
|
offset = 0;
|
||||||
|
|
||||||
|
d3d12_descriptor_heap_get_buffer_range(descriptor_heap, &offset, uav_counter_size, &descriptor_heap->uav_counters);
|
||||||
|
d3d12_descriptor_heap_get_buffer_range(descriptor_heap, &offset, offset_buffer_size, &descriptor_heap->ssbo_ranges);
|
||||||
return S_OK;
|
return S_OK;
|
||||||
}
|
}
|
||||||
|
|
||||||
|
@ -5484,9 +5557,11 @@ static void d3d12_descriptor_heap_update_extra_bindings(struct d3d12_descriptor_
|
||||||
switch (flag)
|
switch (flag)
|
||||||
{
|
{
|
||||||
case VKD3D_BINDLESS_SET_EXTRA_UAV_COUNTER_BUFFER:
|
case VKD3D_BINDLESS_SET_EXTRA_UAV_COUNTER_BUFFER:
|
||||||
vk_buffer->buffer = descriptor_heap->uav_counters.vk_buffer;
|
*vk_buffer = descriptor_heap->uav_counters.descriptor;
|
||||||
vk_buffer->offset = 0;
|
break;
|
||||||
vk_buffer->range = VK_WHOLE_SIZE;
|
|
||||||
|
case VKD3D_BINDLESS_SET_EXTRA_SSBO_OFFSET_BUFFER:
|
||||||
|
*vk_buffer = descriptor_heap->ssbo_ranges.descriptor;
|
||||||
break;
|
break;
|
||||||
|
|
||||||
default:
|
default:
|
||||||
|
@ -5537,21 +5612,8 @@ static HRESULT d3d12_descriptor_heap_init(struct d3d12_descriptor_heap *descript
|
||||||
}
|
}
|
||||||
}
|
}
|
||||||
|
|
||||||
if (desc->Type == D3D12_DESCRIPTOR_HEAP_TYPE_CBV_SRV_UAV &&
|
if (FAILED(hr = d3d12_descriptor_heap_init_data_buffer(descriptor_heap, device, desc)))
|
||||||
(device->bindless_state.flags & VKD3D_RAW_VA_UAV_COUNTER))
|
goto fail;
|
||||||
{
|
|
||||||
if (desc->Flags & D3D12_DESCRIPTOR_HEAP_FLAG_SHADER_VISIBLE)
|
|
||||||
{
|
|
||||||
if (FAILED(hr = d3d12_descriptor_heap_create_uav_counter_buffer(descriptor_heap,
|
|
||||||
&descriptor_heap->uav_counters)))
|
|
||||||
goto fail;
|
|
||||||
}
|
|
||||||
else if (!(descriptor_heap->uav_counters.data = vkd3d_calloc(desc->NumDescriptors, sizeof(VkDeviceSize))))
|
|
||||||
{
|
|
||||||
ERR("Failed to allocate UAV counter address buffer.\n");
|
|
||||||
goto fail;
|
|
||||||
}
|
|
||||||
}
|
|
||||||
|
|
||||||
if (desc->Flags & D3D12_DESCRIPTOR_HEAP_FLAG_SHADER_VISIBLE)
|
if (desc->Flags & D3D12_DESCRIPTOR_HEAP_FLAG_SHADER_VISIBLE)
|
||||||
d3d12_descriptor_heap_update_extra_bindings(descriptor_heap, device);
|
d3d12_descriptor_heap_update_extra_bindings(descriptor_heap, device);
|
||||||
|
@ -5648,11 +5710,11 @@ void d3d12_descriptor_heap_cleanup(struct d3d12_descriptor_heap *descriptor_heap
|
||||||
const struct vkd3d_vk_device_procs *vk_procs = &descriptor_heap->device->vk_procs;
|
const struct vkd3d_vk_device_procs *vk_procs = &descriptor_heap->device->vk_procs;
|
||||||
const struct d3d12_device *device = descriptor_heap->device;
|
const struct d3d12_device *device = descriptor_heap->device;
|
||||||
|
|
||||||
if (!descriptor_heap->uav_counters.vk_memory)
|
if (!descriptor_heap->vk_memory)
|
||||||
vkd3d_free(descriptor_heap->uav_counters.data);
|
vkd3d_free(descriptor_heap->host_memory);
|
||||||
|
|
||||||
VK_CALL(vkDestroyBuffer(device->vk_device, descriptor_heap->uav_counters.vk_buffer, NULL));
|
VK_CALL(vkDestroyBuffer(device->vk_device, descriptor_heap->vk_buffer, NULL));
|
||||||
VK_CALL(vkFreeMemory(device->vk_device, descriptor_heap->uav_counters.vk_memory, NULL));
|
VK_CALL(vkFreeMemory(device->vk_device, descriptor_heap->vk_memory, NULL));
|
||||||
|
|
||||||
VK_CALL(vkDestroyDescriptorPool(device->vk_device, descriptor_heap->vk_descriptor_pool, NULL));
|
VK_CALL(vkDestroyDescriptorPool(device->vk_device, descriptor_heap->vk_descriptor_pool, NULL));
|
||||||
}
|
}
|
||||||
|
|
|
@ -213,7 +213,7 @@ static enum vkd3d_shader_visibility vkd3d_shader_visibility_from_d3d12(D3D12_SHA
|
||||||
|
|
||||||
static VkDescriptorType vk_descriptor_type_from_d3d12_root_parameter(struct d3d12_device *device, D3D12_ROOT_PARAMETER_TYPE type)
|
static VkDescriptorType vk_descriptor_type_from_d3d12_root_parameter(struct d3d12_device *device, D3D12_ROOT_PARAMETER_TYPE type)
|
||||||
{
|
{
|
||||||
bool use_ssbo = d3d12_device_use_ssbo_raw_buffer(device);
|
bool use_ssbo = d3d12_device_use_ssbo_root_descriptors(device);
|
||||||
|
|
||||||
switch (type)
|
switch (type)
|
||||||
{
|
{
|
||||||
|
@ -330,6 +330,7 @@ struct d3d12_root_signature_info
|
||||||
uint32_t push_descriptor_count;
|
uint32_t push_descriptor_count;
|
||||||
uint32_t root_constant_count;
|
uint32_t root_constant_count;
|
||||||
bool has_raw_va_uav_counters;
|
bool has_raw_va_uav_counters;
|
||||||
|
bool has_ssbo_offset_buffer;
|
||||||
|
|
||||||
uint32_t cost;
|
uint32_t cost;
|
||||||
};
|
};
|
||||||
|
@ -345,6 +346,9 @@ static HRESULT d3d12_root_signature_info_count_descriptors(struct d3d12_root_sig
|
||||||
|
|
||||||
if (device->bindless_state.flags & VKD3D_BINDLESS_RAW_SSBO)
|
if (device->bindless_state.flags & VKD3D_BINDLESS_RAW_SSBO)
|
||||||
info->binding_count += 1;
|
info->binding_count += 1;
|
||||||
|
|
||||||
|
if (device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
info->has_ssbo_offset_buffer = true;
|
||||||
break;
|
break;
|
||||||
case D3D12_DESCRIPTOR_RANGE_TYPE_UAV:
|
case D3D12_DESCRIPTOR_RANGE_TYPE_UAV:
|
||||||
/* separate image + buffer descriptors */
|
/* separate image + buffer descriptors */
|
||||||
|
@ -355,6 +359,9 @@ static HRESULT d3d12_root_signature_info_count_descriptors(struct d3d12_root_sig
|
||||||
|
|
||||||
if (device->bindless_state.flags & VKD3D_RAW_VA_UAV_COUNTER)
|
if (device->bindless_state.flags & VKD3D_RAW_VA_UAV_COUNTER)
|
||||||
info->has_raw_va_uav_counters = true;
|
info->has_raw_va_uav_counters = true;
|
||||||
|
|
||||||
|
if (device->bindless_state.flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
info->has_ssbo_offset_buffer = true;
|
||||||
break;
|
break;
|
||||||
case D3D12_DESCRIPTOR_RANGE_TYPE_CBV:
|
case D3D12_DESCRIPTOR_RANGE_TYPE_CBV:
|
||||||
case D3D12_DESCRIPTOR_RANGE_TYPE_SAMPLER:
|
case D3D12_DESCRIPTOR_RANGE_TYPE_SAMPLER:
|
||||||
|
@ -845,6 +852,15 @@ static HRESULT d3d12_root_signature_init(struct d3d12_root_signature *root_signa
|
||||||
&root_signature->uav_counter_binding);
|
&root_signature->uav_counter_binding);
|
||||||
}
|
}
|
||||||
|
|
||||||
|
if (info.has_ssbo_offset_buffer)
|
||||||
|
{
|
||||||
|
root_signature->flags |= VKD3D_ROOT_SIGNATURE_USE_SSBO_OFFSET_BUFFER;
|
||||||
|
|
||||||
|
vkd3d_bindless_state_find_binding(&device->bindless_state,
|
||||||
|
VKD3D_BINDLESS_SET_EXTRA_SSBO_OFFSET_BUFFER,
|
||||||
|
&root_signature->offset_buffer_binding);
|
||||||
|
}
|
||||||
|
|
||||||
if (FAILED(hr = d3d12_root_signature_init_root_descriptors(root_signature, desc,
|
if (FAILED(hr = d3d12_root_signature_init_root_descriptors(root_signature, desc,
|
||||||
&info, &root_signature->push_constant_range, &context,
|
&info, &root_signature->push_constant_range, &context,
|
||||||
&root_signature->vk_root_descriptor_layout)))
|
&root_signature->vk_root_descriptor_layout)))
|
||||||
|
@ -933,6 +949,9 @@ static unsigned int d3d12_root_signature_get_shader_interface_flags(const struct
|
||||||
if (root_signature->flags & VKD3D_ROOT_SIGNATURE_USE_INLINE_UNIFORM_BLOCK)
|
if (root_signature->flags & VKD3D_ROOT_SIGNATURE_USE_INLINE_UNIFORM_BLOCK)
|
||||||
flags |= VKD3D_SHADER_INTERFACE_PUSH_CONSTANTS_AS_UNIFORM_BUFFER;
|
flags |= VKD3D_SHADER_INTERFACE_PUSH_CONSTANTS_AS_UNIFORM_BUFFER;
|
||||||
|
|
||||||
|
if (root_signature->flags & VKD3D_ROOT_SIGNATURE_USE_SSBO_OFFSET_BUFFER)
|
||||||
|
flags |= VKD3D_SHADER_INTERFACE_SSBO_OFFSET_BUFFER;
|
||||||
|
|
||||||
if (root_signature->device->bindless_state.flags & VKD3D_BINDLESS_CBV_AS_SSBO)
|
if (root_signature->device->bindless_state.flags & VKD3D_BINDLESS_CBV_AS_SSBO)
|
||||||
flags |= VKD3D_SHADER_INTERFACE_BINDLESS_CBV_AS_STORAGE_BUFFER;
|
flags |= VKD3D_SHADER_INTERFACE_BINDLESS_CBV_AS_STORAGE_BUFFER;
|
||||||
|
|
||||||
|
@ -1694,6 +1713,7 @@ static HRESULT d3d12_pipeline_state_init_compute(struct d3d12_pipeline_state *st
|
||||||
shader_interface.push_constant_buffers = root_signature->root_constants;
|
shader_interface.push_constant_buffers = root_signature->root_constants;
|
||||||
shader_interface.push_constant_buffer_count = root_signature->root_constant_count;
|
shader_interface.push_constant_buffer_count = root_signature->root_constant_count;
|
||||||
shader_interface.push_constant_ubo_binding = &root_signature->push_constant_ubo_binding;
|
shader_interface.push_constant_ubo_binding = &root_signature->push_constant_ubo_binding;
|
||||||
|
shader_interface.offset_buffer_binding = &root_signature->offset_buffer_binding;
|
||||||
|
|
||||||
if ((hr = vkd3d_create_pipeline_cache_from_d3d12_desc(device, &desc->cached_pso, &state->vk_pso_cache)) < 0)
|
if ((hr = vkd3d_create_pipeline_cache_from_d3d12_desc(device, &desc->cached_pso, &state->vk_pso_cache)) < 0)
|
||||||
{
|
{
|
||||||
|
@ -2527,6 +2547,7 @@ static HRESULT d3d12_pipeline_state_init_graphics(struct d3d12_pipeline_state *s
|
||||||
shader_interface.push_constant_buffers = root_signature->root_constants;
|
shader_interface.push_constant_buffers = root_signature->root_constants;
|
||||||
shader_interface.push_constant_buffer_count = root_signature->root_constant_count;
|
shader_interface.push_constant_buffer_count = root_signature->root_constant_count;
|
||||||
shader_interface.push_constant_ubo_binding = &root_signature->push_constant_ubo_binding;
|
shader_interface.push_constant_ubo_binding = &root_signature->push_constant_ubo_binding;
|
||||||
|
shader_interface.offset_buffer_binding = &root_signature->offset_buffer_binding;
|
||||||
|
|
||||||
graphics->patch_vertex_count = 0;
|
graphics->patch_vertex_count = 0;
|
||||||
|
|
||||||
|
@ -3436,13 +3457,17 @@ static uint32_t vkd3d_bindless_state_get_bindless_flags(struct d3d12_device *dev
|
||||||
|
|
||||||
/* Normally, we would be able to use SSBOs conditionally even when maxSSBOAlignment > 4, but
|
/* Normally, we would be able to use SSBOs conditionally even when maxSSBOAlignment > 4, but
|
||||||
* applications (RE2 being one example) are of course buggy and don't match descriptor and shader usage of resources,
|
* applications (RE2 being one example) are of course buggy and don't match descriptor and shader usage of resources,
|
||||||
* so we cannot rely on alignment analysis to select the appropriate resource type.
|
* so we cannot rely on alignment analysis to select the appropriate resource type. */
|
||||||
* TODO: Implement an offset buffer system so that we can remove the minStorageBufferOffsetAlignment requirement. */
|
|
||||||
if (device_info->descriptor_indexing_properties.maxPerStageDescriptorUpdateAfterBindStorageBuffers >= 1000000 &&
|
if (device_info->descriptor_indexing_properties.maxPerStageDescriptorUpdateAfterBindStorageBuffers >= 1000000 &&
|
||||||
device_info->descriptor_indexing_features.descriptorBindingStorageBufferUpdateAfterBind &&
|
device_info->descriptor_indexing_features.descriptorBindingStorageBufferUpdateAfterBind &&
|
||||||
device_info->properties2.properties.limits.minStorageBufferOffsetAlignment <= 4)
|
device_info->properties2.properties.limits.minStorageBufferOffsetAlignment <= 16)
|
||||||
|
{
|
||||||
flags |= VKD3D_BINDLESS_RAW_SSBO;
|
flags |= VKD3D_BINDLESS_RAW_SSBO;
|
||||||
|
|
||||||
|
if (device_info->properties2.properties.limits.minStorageBufferOffsetAlignment > 4)
|
||||||
|
flags |= VKD3D_SSBO_OFFSET_BUFFER;
|
||||||
|
}
|
||||||
|
|
||||||
if (device_info->buffer_device_address_features.bufferDeviceAddress && (flags & VKD3D_BINDLESS_UAV))
|
if (device_info->buffer_device_address_features.bufferDeviceAddress && (flags & VKD3D_BINDLESS_UAV))
|
||||||
flags |= VKD3D_RAW_VA_UAV_COUNTER;
|
flags |= VKD3D_RAW_VA_UAV_COUNTER;
|
||||||
|
|
||||||
|
@ -3469,6 +3494,9 @@ HRESULT vkd3d_bindless_state_init(struct vkd3d_bindless_state *bindless_state,
|
||||||
if (bindless_state->flags & VKD3D_RAW_VA_UAV_COUNTER)
|
if (bindless_state->flags & VKD3D_RAW_VA_UAV_COUNTER)
|
||||||
extra_bindings |= VKD3D_BINDLESS_SET_EXTRA_UAV_COUNTER_BUFFER;
|
extra_bindings |= VKD3D_BINDLESS_SET_EXTRA_UAV_COUNTER_BUFFER;
|
||||||
|
|
||||||
|
if (bindless_state->flags & VKD3D_SSBO_OFFSET_BUFFER)
|
||||||
|
extra_bindings |= VKD3D_BINDLESS_SET_EXTRA_SSBO_OFFSET_BUFFER;
|
||||||
|
|
||||||
if (FAILED(hr = vkd3d_bindless_state_add_binding(bindless_state, device,
|
if (FAILED(hr = vkd3d_bindless_state_add_binding(bindless_state, device,
|
||||||
VKD3D_BINDLESS_SET_SAMPLER, VK_DESCRIPTOR_TYPE_SAMPLER)))
|
VKD3D_BINDLESS_SET_SAMPLER, VK_DESCRIPTOR_TYPE_SAMPLER)))
|
||||||
goto fail;
|
goto fail;
|
||||||
|
|
|
@ -629,6 +629,7 @@ enum vkd3d_descriptor_flag
|
||||||
VKD3D_DESCRIPTOR_FLAG_DEFINED = (1 << 0),
|
VKD3D_DESCRIPTOR_FLAG_DEFINED = (1 << 0),
|
||||||
VKD3D_DESCRIPTOR_FLAG_VIEW = (1 << 1),
|
VKD3D_DESCRIPTOR_FLAG_VIEW = (1 << 1),
|
||||||
VKD3D_DESCRIPTOR_FLAG_UAV_COUNTER = (1 << 2),
|
VKD3D_DESCRIPTOR_FLAG_UAV_COUNTER = (1 << 2),
|
||||||
|
VKD3D_DESCRIPTOR_FLAG_SSBO_OFFSET = (1 << 3),
|
||||||
};
|
};
|
||||||
|
|
||||||
struct vkd3d_descriptor_binding
|
struct vkd3d_descriptor_binding
|
||||||
|
@ -725,11 +726,16 @@ static inline struct d3d12_dsv_desc *d3d12_dsv_desc_from_cpu_handle(D3D12_CPU_DE
|
||||||
void d3d12_dsv_desc_create_dsv(struct d3d12_dsv_desc *dsv_desc, struct d3d12_device *device,
|
void d3d12_dsv_desc_create_dsv(struct d3d12_dsv_desc *dsv_desc, struct d3d12_device *device,
|
||||||
struct d3d12_resource *resource, const D3D12_DEPTH_STENCIL_VIEW_DESC *desc);
|
struct d3d12_resource *resource, const D3D12_DEPTH_STENCIL_VIEW_DESC *desc);
|
||||||
|
|
||||||
struct d3d12_descriptor_heap_uav_counters
|
struct vkd3d_bound_ssbo_range
|
||||||
{
|
{
|
||||||
VkDeviceAddress *data;
|
uint32_t offset; /* offset to first byte */
|
||||||
VkDeviceMemory vk_memory;
|
uint32_t length; /* bound size in bytes */
|
||||||
VkBuffer vk_buffer;
|
};
|
||||||
|
|
||||||
|
struct vkd3d_host_visible_buffer_range
|
||||||
|
{
|
||||||
|
VkDescriptorBufferInfo descriptor;
|
||||||
|
void *host_ptr;
|
||||||
};
|
};
|
||||||
|
|
||||||
/* ID3D12DescriptorHeap */
|
/* ID3D12DescriptorHeap */
|
||||||
|
@ -743,7 +749,13 @@ struct d3d12_descriptor_heap
|
||||||
VkDescriptorPool vk_descriptor_pool;
|
VkDescriptorPool vk_descriptor_pool;
|
||||||
VkDescriptorSet vk_descriptor_sets[VKD3D_MAX_BINDLESS_DESCRIPTOR_SETS];
|
VkDescriptorSet vk_descriptor_sets[VKD3D_MAX_BINDLESS_DESCRIPTOR_SETS];
|
||||||
|
|
||||||
struct d3d12_descriptor_heap_uav_counters uav_counters;
|
VkDeviceMemory vk_memory;
|
||||||
|
VkBuffer vk_buffer;
|
||||||
|
void *host_memory;
|
||||||
|
|
||||||
|
struct vkd3d_host_visible_buffer_range uav_counters;
|
||||||
|
struct vkd3d_host_visible_buffer_range ssbo_ranges;
|
||||||
|
|
||||||
struct d3d12_device *device;
|
struct d3d12_device *device;
|
||||||
|
|
||||||
struct vkd3d_private_store private_store;
|
struct vkd3d_private_store private_store;
|
||||||
|
@ -804,6 +816,7 @@ enum vkd3d_root_signature_flag
|
||||||
VKD3D_ROOT_SIGNATURE_USE_PUSH_DESCRIPTORS = 0x00000001u,
|
VKD3D_ROOT_SIGNATURE_USE_PUSH_DESCRIPTORS = 0x00000001u,
|
||||||
VKD3D_ROOT_SIGNATURE_USE_INLINE_UNIFORM_BLOCK = 0x00000002u,
|
VKD3D_ROOT_SIGNATURE_USE_INLINE_UNIFORM_BLOCK = 0x00000002u,
|
||||||
VKD3D_ROOT_SIGNATURE_USE_RAW_VA_UAV_COUNTERS = 0x00000004u,
|
VKD3D_ROOT_SIGNATURE_USE_RAW_VA_UAV_COUNTERS = 0x00000004u,
|
||||||
|
VKD3D_ROOT_SIGNATURE_USE_SSBO_OFFSET_BUFFER = 0x00000008u,
|
||||||
};
|
};
|
||||||
|
|
||||||
struct d3d12_root_descriptor_table
|
struct d3d12_root_descriptor_table
|
||||||
|
@ -872,6 +885,7 @@ struct d3d12_root_signature
|
||||||
VkPushConstantRange push_constant_range;
|
VkPushConstantRange push_constant_range;
|
||||||
struct vkd3d_shader_descriptor_binding push_constant_ubo_binding;
|
struct vkd3d_shader_descriptor_binding push_constant_ubo_binding;
|
||||||
struct vkd3d_shader_descriptor_binding uav_counter_binding;
|
struct vkd3d_shader_descriptor_binding uav_counter_binding;
|
||||||
|
struct vkd3d_shader_descriptor_binding offset_buffer_binding;
|
||||||
|
|
||||||
uint32_t descriptor_table_offset;
|
uint32_t descriptor_table_offset;
|
||||||
uint32_t descriptor_table_count;
|
uint32_t descriptor_table_count;
|
||||||
|
@ -1574,6 +1588,7 @@ enum vkd3d_bindless_flags
|
||||||
VKD3D_RAW_VA_UAV_COUNTER = (1u << 4),
|
VKD3D_RAW_VA_UAV_COUNTER = (1u << 4),
|
||||||
VKD3D_BINDLESS_CBV_AS_SSBO = (1u << 5),
|
VKD3D_BINDLESS_CBV_AS_SSBO = (1u << 5),
|
||||||
VKD3D_BINDLESS_RAW_SSBO = (1u << 6),
|
VKD3D_BINDLESS_RAW_SSBO = (1u << 6),
|
||||||
|
VKD3D_SSBO_OFFSET_BUFFER = (1u << 7),
|
||||||
};
|
};
|
||||||
|
|
||||||
#define VKD3D_BINDLESS_SET_MAX_EXTRA_BINDINGS 8
|
#define VKD3D_BINDLESS_SET_MAX_EXTRA_BINDINGS 8
|
||||||
|
@ -1590,6 +1605,7 @@ enum vkd3d_bindless_set_flag
|
||||||
VKD3D_BINDLESS_SET_RAW_SSBO = (1u << 7),
|
VKD3D_BINDLESS_SET_RAW_SSBO = (1u << 7),
|
||||||
|
|
||||||
VKD3D_BINDLESS_SET_EXTRA_UAV_COUNTER_BUFFER = (1u << 24),
|
VKD3D_BINDLESS_SET_EXTRA_UAV_COUNTER_BUFFER = (1u << 24),
|
||||||
|
VKD3D_BINDLESS_SET_EXTRA_SSBO_OFFSET_BUFFER = (1u << 25),
|
||||||
VKD3D_BINDLESS_SET_EXTRA_MASK = 0xff000000u
|
VKD3D_BINDLESS_SET_EXTRA_MASK = 0xff000000u
|
||||||
};
|
};
|
||||||
|
|
||||||
|
@ -1985,6 +2001,14 @@ static inline VkDeviceSize d3d12_device_get_ssbo_alignment(struct d3d12_device *
|
||||||
return device->device_info.properties2.properties.limits.minStorageBufferOffsetAlignment;
|
return device->device_info.properties2.properties.limits.minStorageBufferOffsetAlignment;
|
||||||
}
|
}
|
||||||
|
|
||||||
|
static inline bool d3d12_device_use_ssbo_root_descriptors(struct d3d12_device *device)
|
||||||
|
{
|
||||||
|
/* We only know the VA of root SRV/UAVs, so we cannot
|
||||||
|
* make any better assumptions about the alignment */
|
||||||
|
return d3d12_device_use_ssbo_raw_buffer(device) &&
|
||||||
|
d3d12_device_get_ssbo_alignment(device) <= 4;
|
||||||
|
}
|
||||||
|
|
||||||
/* ID3DBlob */
|
/* ID3DBlob */
|
||||||
struct d3d_blob
|
struct d3d_blob
|
||||||
{
|
{
|
||||||
|
|
|
@ -1 +1 @@
|
||||||
Subproject commit 64ede36e58ebecce64b2f44b908503a8e4ecea19
|
Subproject commit 1f85c7bff8cf6b5defc00e300e1d4540d81ac8d1
|
Loading…
Reference in New Issue