AuroraMiddleware/SPIRV-Cross-Vulnerable
1
0
Fork 0
mirror of https://github.com/KhronosGroup/SPIRV-Cross.git synced 2024-11-14 16:01:07 +00:00
Code Issues Projects Releases Wiki Activity

Merge pull request #571 from KhronosGroup/fix-570

Browse Source 
Handle inout properly with split access chains.
This commit is contained in:
Hans-Kristian Arntzen 2018-05-11 10:39:54 +02:00 committed by GitHub
commit 9d370aca58
No known key found for this signature in database
GPG Key ID: 4AEE18F83AFDEB23
11 changed files with 201 additions and 55 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

47
reference/opt/shaders-msl/comp/struct-packing.comp
View File

@ -97,29 +97,30 @@ struct SSBO0
kernel void main0(device SSBO0& ssbo_140 [[buffer(0)]], device SSBO1& ssbo_430 [[buffer(1)]])
{
ssbo_430.content.m0s[0].a[0] = ssbo_140.content.m0s[0].a[0];
ssbo_430.content.m0s[0].b = ssbo_140.content.m0s[0].b;
ssbo_430.content.m1s[0].a = ssbo_140.content.m1s[0].a;
ssbo_430.content.m1s[0].b = ssbo_140.content.m1s[0].b;
ssbo_430.content.m2s[0].a[0] = ssbo_140.content.m2s[0].a[0];
ssbo_430.content.m2s[0].b = ssbo_140.content.m2s[0].b;
ssbo_430.content.m0.a[0] = ssbo_140.content.m0.a[0];
ssbo_430.content.m0.b = ssbo_140.content.m0.b;
ssbo_430.content.m1.a = ssbo_140.content.m1.a;
ssbo_430.content.m1.b = ssbo_140.content.m1.b;
ssbo_430.content.m2.a[0] = ssbo_140.content.m2.a[0];
ssbo_430.content.m2.b = ssbo_140.content.m2.b;
ssbo_430.content.m3.a = ssbo_140.content.m3.a;
ssbo_430.content.m3.b = ssbo_140.content.m3.b;
ssbo_430.content.m4 = ssbo_140.content.m4;
ssbo_430.content.m3s[0].c = ssbo_140.content.m3s[0].c;
ssbo_430.content.m3s[1].c = ssbo_140.content.m3s[1].c;
ssbo_430.content.m3s[2].c = ssbo_140.content.m3s[2].c;
ssbo_430.content.m3s[3].c = ssbo_140.content.m3s[3].c;
ssbo_430.content.m3s[4].c = ssbo_140.content.m3s[4].c;
ssbo_430.content.m3s[5].c = ssbo_140.content.m3s[5].c;
ssbo_430.content.m3s[6].c = ssbo_140.content.m3s[6].c;
ssbo_430.content.m3s[7].c = ssbo_140.content.m3s[7].c;
Content_1 _60 = ssbo_140.content;
ssbo_430.content.m0s[0].a[0] = _60.m0s[0].a[0];
ssbo_430.content.m0s[0].b = _60.m0s[0].b;
ssbo_430.content.m1s[0].a = _60.m1s[0].a;
ssbo_430.content.m1s[0].b = _60.m1s[0].b;
ssbo_430.content.m2s[0].a[0] = _60.m2s[0].a[0];
ssbo_430.content.m2s[0].b = _60.m2s[0].b;
ssbo_430.content.m0.a[0] = _60.m0.a[0];
ssbo_430.content.m0.b = _60.m0.b;
ssbo_430.content.m1.a = _60.m1.a;
ssbo_430.content.m1.b = _60.m1.b;
ssbo_430.content.m2.a[0] = _60.m2.a[0];
ssbo_430.content.m2.b = _60.m2.b;
ssbo_430.content.m3.a = _60.m3.a;
ssbo_430.content.m3.b = _60.m3.b;
ssbo_430.content.m4 = _60.m4;
ssbo_430.content.m3s[0].c = _60.m3s[0].c;
ssbo_430.content.m3s[1].c = _60.m3s[1].c;
ssbo_430.content.m3s[2].c = _60.m3s[2].c;
ssbo_430.content.m3s[3].c = _60.m3s[3].c;
ssbo_430.content.m3s[4].c = _60.m3s[4].c;
ssbo_430.content.m3s[5].c = _60.m3s[5].c;
ssbo_430.content.m3s[6].c = _60.m3s[6].c;
ssbo_430.content.m3s[7].c = _60.m3s[7].c;
ssbo_430.content.m1.a = ssbo_430.content.m3.a * ssbo_430.m6[1][1];
}

9
reference/opt/shaders/asm/geom/inout-split-access-chain-handle.asm.geom Normal file
View File

@ -0,0 +1,9 @@
#version 440
layout(triangles) in;
layout(max_vertices = 5, triangle_strip) out;
void main()
{
gl_Position = gl_in[0].gl_Position;
}

4
reference/opt/shaders/comp/struct-packing.comp
View File

@ -62,7 +62,7 @@ struct Content_1
S4 m3s[8];
};
layout(binding = 1, std430) buffer SSBO1
layout(binding = 1, std430) restrict buffer SSBO1
{
Content content;
Content content1[2];
@ -78,7 +78,7 @@ layout(binding = 1, std430) buffer SSBO1
float array[];
} ssbo_430;
layout(binding = 0, std140) buffer SSBO0
layout(binding = 0, std140) restrict buffer SSBO0
{
Content_1 content;
Content_1 content1[2];

47
reference/shaders-msl/comp/struct-packing.comp
View File

@ -120,29 +120,30 @@ struct SSBO0
kernel void main0(device SSBO0& ssbo_140 [[buffer(0)]], device SSBO1& ssbo_430 [[buffer(1)]])
{
ssbo_430.content.m0s[0].a[0] = ssbo_140.content.m0s[0].a[0];
ssbo_430.content.m0s[0].b = ssbo_140.content.m0s[0].b;
ssbo_430.content.m1s[0].a = ssbo_140.content.m1s[0].a;
ssbo_430.content.m1s[0].b = ssbo_140.content.m1s[0].b;
ssbo_430.content.m2s[0].a[0] = ssbo_140.content.m2s[0].a[0];
ssbo_430.content.m2s[0].b = ssbo_140.content.m2s[0].b;
ssbo_430.content.m0.a[0] = ssbo_140.content.m0.a[0];
ssbo_430.content.m0.b = ssbo_140.content.m0.b;
ssbo_430.content.m1.a = ssbo_140.content.m1.a;
ssbo_430.content.m1.b = ssbo_140.content.m1.b;
ssbo_430.content.m2.a[0] = ssbo_140.content.m2.a[0];
ssbo_430.content.m2.b = ssbo_140.content.m2.b;
ssbo_430.content.m3.a = ssbo_140.content.m3.a;
ssbo_430.content.m3.b = ssbo_140.content.m3.b;
ssbo_430.content.m4 = ssbo_140.content.m4;
ssbo_430.content.m3s[0].c = ssbo_140.content.m3s[0].c;
ssbo_430.content.m3s[1].c = ssbo_140.content.m3s[1].c;
ssbo_430.content.m3s[2].c = ssbo_140.content.m3s[2].c;
ssbo_430.content.m3s[3].c = ssbo_140.content.m3s[3].c;
ssbo_430.content.m3s[4].c = ssbo_140.content.m3s[4].c;
ssbo_430.content.m3s[5].c = ssbo_140.content.m3s[5].c;
ssbo_430.content.m3s[6].c = ssbo_140.content.m3s[6].c;
ssbo_430.content.m3s[7].c = ssbo_140.content.m3s[7].c;
Content_1 _60 = ssbo_140.content;
ssbo_430.content.m0s[0].a[0] = _60.m0s[0].a[0];
ssbo_430.content.m0s[0].b = _60.m0s[0].b;
ssbo_430.content.m1s[0].a = _60.m1s[0].a;
ssbo_430.content.m1s[0].b = _60.m1s[0].b;
ssbo_430.content.m2s[0].a[0] = _60.m2s[0].a[0];
ssbo_430.content.m2s[0].b = _60.m2s[0].b;
ssbo_430.content.m0.a[0] = _60.m0.a[0];
ssbo_430.content.m0.b = _60.m0.b;
ssbo_430.content.m1.a = _60.m1.a;
ssbo_430.content.m1.b = _60.m1.b;
ssbo_430.content.m2.a[0] = _60.m2.a[0];
ssbo_430.content.m2.b = _60.m2.b;
ssbo_430.content.m3.a = _60.m3.a;
ssbo_430.content.m3.b = _60.m3.b;
ssbo_430.content.m4 = _60.m4;
ssbo_430.content.m3s[0].c = _60.m3s[0].c;
ssbo_430.content.m3s[1].c = _60.m3s[1].c;
ssbo_430.content.m3s[2].c = _60.m3s[2].c;
ssbo_430.content.m3s[3].c = _60.m3s[3].c;
ssbo_430.content.m3s[4].c = _60.m3s[4].c;
ssbo_430.content.m3s[5].c = _60.m3s[5].c;
ssbo_430.content.m3s[6].c = _60.m3s[6].c;
ssbo_430.content.m3s[7].c = _60.m3s[7].c;
ssbo_430.content.m1.a = ssbo_430.content.m3.a * ssbo_430.m6[1][1];
}

23
reference/shaders/asm/geom/inout-split-access-chain-handle.asm.geom Normal file
View File

@ -0,0 +1,23 @@
#version 440
layout(triangles) in;
layout(max_vertices = 5, triangle_strip) out;
struct Data
{
vec4 ApiPerspectivePosition;
};
void Copy(inout Data inputStream[3])
{
inputStream[0].ApiPerspectivePosition = gl_in[0].gl_Position;
}
void main()
{
Data inputStream[3];
Data param[3] = inputStream;
Copy(param);
inputStream = param;
gl_Position = inputStream[0].ApiPerspectivePosition;
}

4
reference/shaders/comp/struct-packing.comp
View File

@ -85,7 +85,7 @@ struct Content_1
S4_1 m3s[8];
};
layout(binding = 1, std430) buffer SSBO1
layout(binding = 1, std430) restrict buffer SSBO1
{
Content content;
Content content1[2];
@ -101,7 +101,7 @@ layout(binding = 1, std430) buffer SSBO1
float array[];
} ssbo_430;
layout(binding = 0, std140) buffer SSBO0
layout(binding = 0, std140) restrict buffer SSBO0
{
Content_1 content;
Content_1 content1[2];

90
shaders/asm/geom/inout-split-access-chain-handle.asm.geom Normal file
View File

@ -0,0 +1,90 @@
; SPIR-V
; Version: 1.0
; Generator: Khronos Glslang Reference Front End; 3
; Bound: 42
; Schema: 0
OpCapability Geometry
%1 = OpExtInstImport "GLSL.std.450"
OpMemoryModel Logical GLSL450
OpEntryPoint Geometry %main "main" %gl_in %_
OpExecutionMode %main Triangles
OpExecutionMode %main Invocations 1
OpExecutionMode %main OutputTriangleStrip
OpExecutionMode %main OutputVertices 5
OpSource GLSL 440
OpName %main "main"
OpName %Data "Data"
OpMemberName %Data 0 "ApiPerspectivePosition"
OpName %Copy_struct_Data_vf41_3__ "Copy(struct-Data-vf41[3];"
OpName %inputStream "inputStream"
OpName %gl_PerVertex "gl_PerVertex"
OpMemberName %gl_PerVertex 0 "gl_Position"
OpMemberName %gl_PerVertex 1 "gl_PointSize"
OpMemberName %gl_PerVertex 2 "gl_ClipDistance"
OpName %gl_in "gl_in"
OpName %inputStream_0 "inputStream"
OpName %param "param"
OpName %gl_PerVertex_0 "gl_PerVertex"
OpMemberName %gl_PerVertex_0 0 "gl_Position"
OpMemberName %gl_PerVertex_0 1 "gl_PointSize"
OpMemberName %gl_PerVertex_0 2 "gl_ClipDistance"
OpName %_ ""
OpMemberDecorate %gl_PerVertex 0 BuiltIn Position
OpMemberDecorate %gl_PerVertex 1 BuiltIn PointSize
OpMemberDecorate %gl_PerVertex 2 BuiltIn ClipDistance
OpDecorate %gl_PerVertex Block
OpMemberDecorate %gl_PerVertex_0 0 BuiltIn Position
OpMemberDecorate %gl_PerVertex_0 1 BuiltIn PointSize
OpMemberDecorate %gl_PerVertex_0 2 BuiltIn ClipDistance
OpDecorate %gl_PerVertex_0 Block
%void = OpTypeVoid
%3 = OpTypeFunction %void
%float = OpTypeFloat 32
%v4float = OpTypeVector %float 4
%Data = OpTypeStruct %v4float
%uint = OpTypeInt 32 0
%uint_3 = OpConstant %uint 3
%_arr_Data_uint_3 = OpTypeArray %Data %uint_3
%_ptr_Function__Data = OpTypePointer Function %Data
%_ptr_Function__arr_Data_uint_3 = OpTypePointer Function %_arr_Data_uint_3
%13 = OpTypeFunction %void %_ptr_Function__arr_Data_uint_3
%int = OpTypeInt 32 1
%int_0 = OpConstant %int 0
%uint_1 = OpConstant %uint 1
%_arr_float_uint_1 = OpTypeArray %float %uint_1
%gl_PerVertex = OpTypeStruct %v4float %float %_arr_float_uint_1
%_arr_gl_PerVertex_uint_3 = OpTypeArray %gl_PerVertex %uint_3
%_ptr_Input__arr_gl_PerVertex_uint_3 = OpTypePointer Input %_arr_gl_PerVertex_uint_3
%gl_in = OpVariable %_ptr_Input__arr_gl_PerVertex_uint_3 Input
%_ptr_Input_v4float = OpTypePointer Input %v4float
%_ptr_Function_v4float = OpTypePointer Function %v4float
%gl_PerVertex_0 = OpTypeStruct %v4float %float %_arr_float_uint_1
%_ptr_Output_gl_PerVertex_0 = OpTypePointer Output %gl_PerVertex_0
%_ = OpVariable %_ptr_Output_gl_PerVertex_0 Output
%_ptr_Output_v4float = OpTypePointer Output %v4float
%main = OpFunction %void None %3
%5 = OpLabel
%inputStream_0 = OpVariable %_ptr_Function__arr_Data_uint_3 Function
%param = OpVariable %_ptr_Function__arr_Data_uint_3 Function
%32 = OpLoad %_arr_Data_uint_3 %inputStream_0
OpStore %param %32
%33 = OpFunctionCall %void %Copy_struct_Data_vf41_3__ %param
%34 = OpLoad %_arr_Data_uint_3 %param
OpStore %inputStream_0 %34
%59 = OpAccessChain %_ptr_Function__Data %inputStream_0 %int_0
%38 = OpAccessChain %_ptr_Function_v4float %59 %int_0
%39 = OpLoad %v4float %38
%41 = OpAccessChain %_ptr_Output_v4float %_ %int_0
OpStore %41 %39
OpReturn
OpFunctionEnd
%Copy_struct_Data_vf41_3__ = OpFunction %void None %13
%inputStream = OpFunctionParameter %_ptr_Function__arr_Data_uint_3
%16 = OpLabel
%26 = OpAccessChain %_ptr_Input_v4float %gl_in %int_0 %int_0
%27 = OpLoad %v4float %26
%28 = OpAccessChain %_ptr_Function__Data %inputStream %int_0
%29 = OpAccessChain %_ptr_Function_v4float %28 %int_0
OpStore %29 %27
OpReturn
OpFunctionEnd

4
shaders/comp/struct-packing.comp
View File

@ -44,7 +44,7 @@ struct Content
S4 m3s[8];
};
layout(binding = 1, std430) buffer SSBO1
layout(binding = 1, std430) restrict buffer SSBO1
{
Content content;
Content content1[2];
@ -61,7 +61,7 @@ layout(binding = 1, std430) buffer SSBO1
float array[];
} ssbo_430;
layout(binding = 0, std140) buffer SSBO0
layout(binding = 0, std140) restrict buffer SSBO0
{
Content content;
Content content1[2];

9
spirv_common.hpp
View File

@ -1111,9 +1111,10 @@ public:
void set(std::unique_ptr<IVariant> val, uint32_t new_type)
{
holder = std::move(val);
if (type != TypeNone && type != new_type)
if (!allow_type_rewrite && type != TypeNone && type != new_type)
SPIRV_CROSS_THROW("Overwriting a variant with new type.");
type = new_type;
allow_type_rewrite = false;
}
template <typename T>
@ -1154,9 +1155,15 @@ public:
type = TypeNone;
}
void set_allow_type_rewrite()
{
allow_type_rewrite = true;
}
private:
std::unique_ptr<IVariant> holder;
uint32_t type = TypeNone;
bool allow_type_rewrite = false;
};
template <typename T>

15
spirv_cross.cpp
View File

@ -104,7 +104,13 @@ bool Compiler::variable_storage_is_aliased(const SPIRVariable &v)
meta[type.self].decoration.decoration_flags.get(DecorationBufferBlock);
bool image = type.basetype == SPIRType::Image;
bool counter = type.basetype == SPIRType::AtomicCounter;
bool is_restrict = meta[v.self].decoration.decoration_flags.get(DecorationRestrict);
bool is_restrict;
if (ssbo)
is_restrict = get_buffer_block_flags(v).get(DecorationRestrict);
else
is_restrict = has_decoration(v.self, DecorationRestrict);
return !is_restrict && (ssbo || image || counter);
}
@ -3191,6 +3197,9 @@ bool Compiler::DummySamplerForCombinedImageHandler::handle(Op opcode, const uint
uint32_t ptr = args[2];
compiler.set<SPIRExpression>(id, "", result_type, true);
compiler.register_read(id, ptr, true);
// Other backends might use SPIRAccessChain for this later.
compiler.ids[id].set_allow_type_rewrite();
break;
}
@ -3696,6 +3705,10 @@ void Compiler::analyze_variable_scope(SPIRFunction &entry)
notify_variable_access(args[i], current_block->self);
// The result of an access chain is a fixed expression and is not really considered a temporary.
auto &e = compiler.set<SPIRExpression>(args[1], "", args[0], true);
auto *backing_variable = compiler.maybe_get_backing_variable(ptr);
e.loaded_from = backing_variable ? backing_variable->self : 0;
compiler.ids[args[1]].set_allow_type_rewrite();
break;
}

4
spirv_glsl.cpp
View File

@ -6073,7 +6073,9 @@ void CompilerGLSL::emit_instruction(const Instruction &instruction)
bool need_transpose, result_is_packed;
auto e = access_chain(ops[2], &ops[3], length - 3, get<SPIRType>(ops[0]), &need_transpose, &result_is_packed);
auto &expr = set<SPIRExpression>(ops[1], move(e), ops[0], should_forward(ops[2]));
expr.loaded_from = ops[2];
auto *backing_variable = maybe_get_backing_variable(ops[2]);
expr.loaded_from = backing_variable ? backing_variable->self : ops[2];
expr.need_transpose = need_transpose;
// Mark the result as being packed. Some platforms handled packed vectors differently than non-packed.