shader_info: Move subgroup_size out of cs and make it an enum
Reviewed-by: Caio Oliveira <caio.oliveira@intel.com> Part-of: <https://gitlab.freedesktop.org/mesa/mesa/-/merge_requests/17337>
This commit is contained in:
parent
e64fd5e475
commit
e1ee201722
|
@ -26,6 +26,8 @@
|
||||||
#ifndef SHADER_ENUMS_H
|
#ifndef SHADER_ENUMS_H
|
||||||
#define SHADER_ENUMS_H
|
#define SHADER_ENUMS_H
|
||||||
|
|
||||||
|
#include "util/macros.h"
|
||||||
|
|
||||||
#include <stdbool.h>
|
#include <stdbool.h>
|
||||||
|
|
||||||
/* Project-wide (GL and Vulkan) maximum. */
|
/* Project-wide (GL and Vulkan) maximum. */
|
||||||
|
@ -1215,6 +1217,21 @@ enum cl_sampler_filter_mode {
|
||||||
#define MAT_BIT_FRONT_INDEXES (1<<MAT_ATTRIB_FRONT_INDEXES)
|
#define MAT_BIT_FRONT_INDEXES (1<<MAT_ATTRIB_FRONT_INDEXES)
|
||||||
#define MAT_BIT_BACK_INDEXES (1<<MAT_ATTRIB_BACK_INDEXES)
|
#define MAT_BIT_BACK_INDEXES (1<<MAT_ATTRIB_BACK_INDEXES)
|
||||||
|
|
||||||
|
/** An enum representing what kind of input gl_SubgroupSize is. */
|
||||||
|
enum PACKED gl_subgroup_size
|
||||||
|
{
|
||||||
|
/** Actual subgroup size, whatever that happens to be */
|
||||||
|
SUBGROUP_SIZE_VARYING = 0,
|
||||||
|
|
||||||
|
/* These enums are specifically chosen so that the value of the enum is
|
||||||
|
* also the subgroup size. If any new values are added, they must respect
|
||||||
|
* this invariant.
|
||||||
|
*/
|
||||||
|
SUBGROUP_SIZE_REQUIRE_8 = 8, /**< VK_EXT_subgroup_size_control */
|
||||||
|
SUBGROUP_SIZE_REQUIRE_16 = 16, /**< VK_EXT_subgroup_size_control */
|
||||||
|
SUBGROUP_SIZE_REQUIRE_32 = 32, /**< VK_EXT_subgroup_size_control */
|
||||||
|
};
|
||||||
|
|
||||||
#ifdef __cplusplus
|
#ifdef __cplusplus
|
||||||
} /* extern "C" */
|
} /* extern "C" */
|
||||||
#endif
|
#endif
|
||||||
|
|
|
@ -237,6 +237,8 @@ typedef struct shader_info {
|
||||||
*/
|
*/
|
||||||
uint16_t workgroup_size[3];
|
uint16_t workgroup_size[3];
|
||||||
|
|
||||||
|
enum gl_subgroup_size subgroup_size;
|
||||||
|
|
||||||
/* Transform feedback buffer strides in dwords, max. 1K - 4. */
|
/* Transform feedback buffer strides in dwords, max. 1K - 4. */
|
||||||
uint8_t xfb_stride[MAX_XFB_BUFFERS];
|
uint8_t xfb_stride[MAX_XFB_BUFFERS];
|
||||||
|
|
||||||
|
@ -481,11 +483,6 @@ typedef struct shader_info {
|
||||||
*/
|
*/
|
||||||
enum gl_derivative_group derivative_group:2;
|
enum gl_derivative_group derivative_group:2;
|
||||||
|
|
||||||
/**
|
|
||||||
* Explicit subgroup size if set by the shader, otherwise 0.
|
|
||||||
*/
|
|
||||||
unsigned subgroup_size;
|
|
||||||
|
|
||||||
/**
|
/**
|
||||||
* pointer size is:
|
* pointer size is:
|
||||||
* AddressingModelLogical: 0 (default)
|
* AddressingModelLogical: 0 (default)
|
||||||
|
|
|
@ -5262,7 +5262,7 @@ vtn_handle_execution_mode(struct vtn_builder *b, struct vtn_value *entry_point,
|
||||||
|
|
||||||
case SpvExecutionModeSubgroupSize:
|
case SpvExecutionModeSubgroupSize:
|
||||||
vtn_assert(b->shader->info.stage == MESA_SHADER_KERNEL);
|
vtn_assert(b->shader->info.stage == MESA_SHADER_KERNEL);
|
||||||
b->shader->info.cs.subgroup_size = mode->operands[0];
|
b->shader->info.subgroup_size = mode->operands[0];
|
||||||
break;
|
break;
|
||||||
|
|
||||||
case SpvExecutionModeSubgroupUniformControlFlowKHR:
|
case SpvExecutionModeSubgroupUniformControlFlowKHR:
|
||||||
|
|
|
@ -41,9 +41,10 @@ brw_required_dispatch_width(const struct shader_info *info,
|
||||||
required = (unsigned)subgroup_size_type;
|
required = (unsigned)subgroup_size_type;
|
||||||
}
|
}
|
||||||
|
|
||||||
if (gl_shader_stage_is_compute(info->stage) && info->cs.subgroup_size > 0) {
|
if (gl_shader_stage_is_compute(info->stage) &&
|
||||||
assert(required == 0 || required == info->cs.subgroup_size);
|
info->subgroup_size >= SUBGROUP_SIZE_REQUIRE_8) {
|
||||||
required = info->cs.subgroup_size;
|
assert(required == 0 || required == info->subgroup_size);
|
||||||
|
required = info->subgroup_size;
|
||||||
}
|
}
|
||||||
|
|
||||||
return required;
|
return required;
|
||||||
|
|
Loading…
Reference in New Issue