Skip to content

Commit b82d94e

Browse files
WebGL infinite loops are optimized out in certain cases
https://bugs.webkit.org/show_bug.cgi?id=281902 rdar://136486349 Reviewed by Mike Wyrzykowski. Metal: Ensure potentially infinite loops have defined behavior The MSL compiler would omit infinite loops and assume number domains based on the omission logic. This would induce incorrect number domains in case the infinite loops would be invokable. Infinite loops are undefined in C++ and thus in MSL. It is the job of the programmer to ensure undefined behavior cannot happen. Consider GLSL loop like: uniform float i; ... if (i != 0.5) for(;;) { } gl_FragColor = vec4(i); Historically this would emit MSL loop in spirit of: if (i != 0.5) { bool c = true; while (c) { } } ANGLE_fragmentOut.gl_FragColor = metal::float4(i, i, i, i); Since This could cause the MSL compiler to optimize the function to equivalent of: ANGLE_fragmentOut.gl_FragColor = metal::float4(0.5, 0.5, 0.5, 0.5); Presumably this loop omission would happen at the clang frontend part. Before, was worked around by emitting asm statements to the MSL: bool c = true; while (c) { __asm__(""); } The asm injection would would work for this particular source pattern, presumably because injecting the asm would avoid the loop omission at the clang frontend part. The MSL/C++ code is still UB, though. The asm statement does not cause anything that C++ would consider as "forward progress" of the loop. The success was just due to how the backend worked. The bitcode produced would be similar to: 4: tail call void asm sideeffect "", ""() #6, !srcloc !28 br label %4, !llvm.loop !29 Here, the compiler can be seen to simply fail to detect a loop that does not make forward progress. Considering GLSL of form: uniform int f; ... for (;;) { if (f <= 1) break; } With asm injection to the loop, this would produce: 5: tail call void asm sideeffect "", ""() #8, !srcloc !29 %6 = load i32, i32 addrspace(2)* %4, align 4, !tbaa !30 %7 = icmp slt i32 %6, 2 br i1 %7, label %8, label %5 8: This code is still assumed to make progress. The backend optimizer is free to assume that the condition holds, since the load to break the loop is from constant address space. I.e. uniform f does not change its value during the loop. Instead of injecting asm, inject a read of unused volatile variable. The volatile variable access is defined in C++ as forward progress. This means infinite loop containing such read is considered defined. To simplify the implementation and to avoid volatile writes, the read is to a dummy variable instead of the loop condition bool. The tests here do not pass completely for MSL backend. In case the compiler would omit the infinite loop (unpatched code), they would fail with demonstration of how the values behave. After fixing, the loops cause timeout but Metal backend does not have implementation to report context loss. Also, the ReadPixels is just for demostration purposes of the unpatched code. * Source/ThirdParty/ANGLE/src/compiler/translator/msl/EmitMetal.cpp: (GenMetalTraverser::GenMetalTraverser): (GenMetalTraverser::emitLoopBody): (GenMetalTraverser::emitForwardProgressStore): (GenMetalTraverser::emitForwardProgressSignal): (GenMetalTraverser::visitForLoop): (GenMetalTraverser::visitWhileLoop): (GenMetalTraverser::visitDoWhileLoop): * Source/ThirdParty/ANGLE/src/tests/angle_end2end_tests.gni: * Source/ThirdParty/ANGLE/src/tests/gl_tests/TimeoutDrawTest.cpp: Added. (angle::TimeoutDrawTest::TimeoutDrawTest): (angle::TEST_P): Canonical link: https://commits.webkit.org/283286.350@safari-7620-branch
1 parent 47ed21f commit b82d94e

File tree

3 files changed

+252
-11
lines changed

3 files changed

+252
-11
lines changed

Source/ThirdParty/ANGLE/src/compiler/translator/msl/EmitMetal.cpp

Lines changed: 75 additions & 11 deletions
Original file line numberDiff line numberDiff line change
@@ -181,12 +181,17 @@ class GenMetalTraverser : public TIntermTraverser
181181

182182
void emitSingleConstant(const TConstantUnion *const constUnion);
183183

184+
void emitForwardProgressStore();
185+
void emitForwardProgressSignal();
186+
bool shouldEnsureForwardProgress() const { return mForwardProgressStoreNestingCount >= 0; }
187+
184188
private:
185189
Sink &mOut;
186190
const TCompiler &mCompiler;
187191
const PipelineStructs &mPipelineStructs;
188192
SymbolEnv &mSymbolEnv;
189193
IdGen &mIdGen;
194+
int mForwardProgressStoreNestingCount = -1; // Negative means forward progress is not ensured.
190195
int mIndentLevel = -1;
191196
int mLastIndentationPos = -1;
192197
int mOpenPointerParenCount = 0;
@@ -201,8 +206,44 @@ class GenMetalTraverser : public TIntermTraverser
201206
size_t mDriverUniformsBindingIndex = 0;
202207
size_t mUBOArgumentBufferBindingIndex = 0;
203208
bool mRasterOrderGroupsSupported = false;
204-
bool mInjectAsmStatementIntoLoopBodies = false;
209+
friend class ScopedForwardProgressStore;
205210
};
211+
212+
class ScopedForwardProgressStore
213+
{
214+
public:
215+
ScopedForwardProgressStore(GenMetalTraverser& traverser);
216+
~ScopedForwardProgressStore();
217+
private:
218+
GenMetalTraverser& mTraverser;
219+
};
220+
221+
ScopedForwardProgressStore::ScopedForwardProgressStore(GenMetalTraverser& traverser)
222+
: mTraverser(traverser)
223+
{
224+
if (mTraverser.shouldEnsureForwardProgress())
225+
{
226+
if (mTraverser.mForwardProgressStoreNestingCount == 0)
227+
{
228+
mTraverser.emitOpenBrace();
229+
mTraverser.emitForwardProgressStore();
230+
}
231+
++mTraverser.mForwardProgressStoreNestingCount;
232+
}
233+
}
234+
235+
ScopedForwardProgressStore::~ScopedForwardProgressStore()
236+
{
237+
if (mTraverser.shouldEnsureForwardProgress())
238+
{
239+
--mTraverser.mForwardProgressStoreNestingCount;
240+
if (mTraverser.mForwardProgressStoreNestingCount == 0)
241+
{
242+
mTraverser.emitCloseBrace();
243+
}
244+
}
245+
}
246+
206247
} // anonymous namespace
207248

208249
GenMetalTraverser::~GenMetalTraverser()
@@ -228,9 +269,11 @@ GenMetalTraverser::GenMetalTraverser(const TCompiler &compiler,
228269
mDriverUniformsBindingIndex(compileOptions.metal.driverUniformsBindingIndex),
229270
mUBOArgumentBufferBindingIndex(compileOptions.metal.UBOArgumentBufferBindingIndex),
230271
mRasterOrderGroupsSupported(compileOptions.pls.fragmentSyncType ==
231-
ShFragmentSynchronizationType::RasterOrderGroups_Metal),
232-
mInjectAsmStatementIntoLoopBodies(compileOptions.metal.injectAsmStatementIntoLoopBodies)
233-
{}
272+
ShFragmentSynchronizationType::RasterOrderGroups_Metal)
273+
{
274+
if (compileOptions.metal.injectAsmStatementIntoLoopBodies)
275+
mForwardProgressStoreNestingCount = 0;
276+
}
234277

235278
void GenMetalTraverser::emitIndentation()
236279
{
@@ -887,17 +930,14 @@ void GenMetalTraverser::emitPostQualifier(const EmitVariableDeclarationConfig &e
887930

888931
void GenMetalTraverser::emitLoopBody(TIntermBlock *bodyNode)
889932
{
890-
if (mInjectAsmStatementIntoLoopBodies)
933+
const bool emitForwardProgress = shouldEnsureForwardProgress();
934+
if (emitForwardProgress)
891935
{
892936
emitOpenBrace();
893-
894-
emitIndentation();
895-
mOut << "__asm__(\"\");\n";
937+
emitForwardProgressSignal();
896938
}
897-
898939
bodyNode->traverse(this);
899-
900-
if (mInjectAsmStatementIntoLoopBodies)
940+
if (emitForwardProgress)
901941
{
902942
emitCloseBrace();
903943
}
@@ -2415,6 +2455,26 @@ void GenMetalTraverser::emitCloseBrace()
24152455
mOut << "}";
24162456
}
24172457

2458+
void GenMetalTraverser::emitForwardProgressStore()
2459+
{
2460+
// https://eel.is/c++draft/intro.progress
2461+
// "The implementation may assume that any thread will eventually do one of the following:""
2462+
// - ...
2463+
// - "perform an access through a volatile glvalue"
2464+
// Emit a volatile variable which all loops in the stack will access.
2465+
emitIndentation();
2466+
mOut << "volatile bool ANGLE_p;\n";
2467+
}
2468+
2469+
void GenMetalTraverser::emitForwardProgressSignal()
2470+
{
2471+
// Emit a read though the volatile variable. This marks the loop as making forward progress even
2472+
// if the compiler can otherwise analyze it to be infinite. This ensures that the loop
2473+
// has defined behavior.
2474+
emitIndentation();
2475+
mOut << "(void) ANGLE_p;\n";
2476+
}
2477+
24182478
static bool RequiresSemicolonTerminator(TIntermNode &node)
24192479
{
24202480
if (node.getAsBlock())
@@ -2616,6 +2676,8 @@ bool GenMetalTraverser::visitForLoop(TIntermLoop *loopNode)
26162676
TIntermTyped *condNode = loopNode->getCondition();
26172677
TIntermTyped *exprNode = loopNode->getExpression();
26182678

2679+
ScopedForwardProgressStore scopedProgress(*this);
2680+
26192681
mOut << "for (";
26202682

26212683
if (initNode)
@@ -2658,6 +2720,7 @@ bool GenMetalTraverser::visitWhileLoop(TIntermLoop *loopNode)
26582720
ASSERT(condNode);
26592721
ASSERT(!initNode && !exprNode);
26602722

2723+
ScopedForwardProgressStore scopedProgress(*this);
26612724
emitIndentation();
26622725
mOut << "while (";
26632726
condNode->traverse(this);
@@ -2677,6 +2740,7 @@ bool GenMetalTraverser::visitDoWhileLoop(TIntermLoop *loopNode)
26772740
ASSERT(condNode);
26782741
ASSERT(!initNode && !exprNode);
26792742

2743+
ScopedForwardProgressStore scopedProgress(*this);
26802744
emitIndentation();
26812745
mOut << "do\n";
26822746
emitLoopBody(loopNode->getBody());

Source/ThirdParty/ANGLE/src/tests/angle_end2end_tests.gni

Lines changed: 1 addition & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -164,6 +164,7 @@ angle_end2end_tests_sources = [
164164
"gl_tests/TextureTest.cpp",
165165
"gl_tests/TextureUploadFormatTest.cpp",
166166
"gl_tests/TiledRenderingTest.cpp",
167+
"gl_tests/TimeoutDrawTest.cpp",
167168
"gl_tests/TimerQueriesTest.cpp",
168169
"gl_tests/TransformFeedbackTest.cpp",
169170
"gl_tests/UniformBufferTest.cpp",
Lines changed: 176 additions & 0 deletions
Original file line numberDiff line numberDiff line change
@@ -0,0 +1,176 @@
1+
//
2+
// Copyright 2015 The ANGLE Project Authors. All rights reserved.
3+
// Use of this source code is governed by a BSD-style license that can be
4+
// found in the LICENSE file.
5+
//
6+
7+
#include "test_utils/ANGLETest.h"
8+
9+
#include "test_utils/gl_raii.h"
10+
#include "util/shader_utils.h"
11+
12+
using namespace angle;
13+
14+
namespace
15+
{
16+
class TimeoutDrawTest : public ANGLETest<>
17+
{
18+
protected:
19+
TimeoutDrawTest()
20+
{
21+
setWindowWidth(128);
22+
setWindowHeight(128);
23+
setConfigRedBits(8);
24+
setConfigGreenBits(8);
25+
setConfigBlueBits(8);
26+
setConfigAlphaBits(8);
27+
// Tests should skip if robustness not supported, but this can be done only after
28+
// Metal supports robustness.
29+
if (IsEGLClientExtensionEnabled("EGL_EXT_create_context_robustness"))
30+
{
31+
setContextResetStrategy(EGL_LOSE_CONTEXT_ON_RESET_EXT);
32+
}
33+
else
34+
{
35+
setContextResetStrategy(EGL_NO_RESET_NOTIFICATION_EXT);
36+
}
37+
}
38+
void testSetUp() override
39+
{
40+
glClear(GL_COLOR_BUFFER_BIT);
41+
glFinish();
42+
}
43+
};
44+
45+
// Tests that trivial infinite loops in vertex shaders hang instead of progress.
46+
TEST_P(TimeoutDrawTest, TrivialInfiniteLoopVS)
47+
{
48+
constexpr char kVS[] = R"(precision highp float;
49+
attribute vec4 a_position;
50+
void main()
51+
{
52+
for (;;) {}
53+
gl_Position = a_position;
54+
})";
55+
ANGLE_GL_PROGRAM(program, kVS, essl1_shaders::fs::Red());
56+
drawQuad(program, essl1_shaders::PositionAttrib(), 0.5f);
57+
glFinish();
58+
EXPECT_GL_ERROR(GL_CONTEXT_LOST);
59+
EXPECT_PIXEL_COLOR_EQ(0, 0, GLColor::transparentBlack); // Should read through client buffer since context should be lost.
60+
}
61+
62+
// Tests that trivial infinite loops in fragment shaders hang instead of progress.
63+
TEST_P(TimeoutDrawTest, TrivialInfiniteLoopFS)
64+
{
65+
constexpr char kFS[] = R"(precision mediump float;
66+
void main()
67+
{
68+
for (;;) {}
69+
gl_FragColor = vec4(1, 0, 0, 1);
70+
})";
71+
ANGLE_GL_PROGRAM(program, essl1_shaders::vs::Simple(), kFS);
72+
drawQuad(program, essl1_shaders::PositionAttrib(), 0.5f);
73+
glFinish();
74+
EXPECT_GL_ERROR(GL_CONTEXT_LOST);
75+
EXPECT_PIXEL_COLOR_EQ(0, 0, GLColor::transparentBlack); // Should read through client buffer since context should be lost.
76+
}
77+
78+
79+
// Tests that infinite loops based on user-supplied values in vertex shaders hang instead of progress.
80+
// Otherwise optimizer would be able to assume something about the domain of the user-supplied value.
81+
TEST_P(TimeoutDrawTest, DynamicInfiniteLoopVS)
82+
{
83+
constexpr char kVS[] = R"(precision highp float;
84+
attribute vec4 a_position;
85+
uniform int f;
86+
void main()
87+
{
88+
for (;f != 0;) {}
89+
gl_Position = a_position;
90+
})";
91+
ANGLE_GL_PROGRAM(program, kVS, essl1_shaders::fs::Red());
92+
93+
glUseProgram(program);
94+
GLint uniformLocation = glGetUniformLocation(program, "f");
95+
glUniform1i(uniformLocation, 77);
96+
97+
drawQuad(program, essl1_shaders::PositionAttrib(), 0.5f);
98+
glFinish();
99+
EXPECT_PIXEL_COLOR_EQ(0, 0, GLColor::transparentBlack); // Should read through client buffer since context should be lost.
100+
EXPECT_GL_ERROR(GL_CONTEXT_LOST);
101+
}
102+
103+
// Tests that infinite loops based on user-supplied values in fragment shaders hang instead of progress.
104+
// Otherwise optimizer would be able to assume something about the domain of the user-supplied value.
105+
TEST_P(TimeoutDrawTest, DynamicInfiniteLoopFS)
106+
{
107+
constexpr char kFS[] = R"(precision mediump float;
108+
uniform int f;
109+
void main()
110+
{
111+
for (;f != 0;) {}
112+
gl_FragColor = vec4(1, 0, 0, 1);
113+
})";
114+
ANGLE_GL_PROGRAM(program, essl1_shaders::vs::Simple(), kFS);
115+
glUseProgram(program);
116+
GLint uniformLocation = glGetUniformLocation(program, "f");
117+
glUniform1i(uniformLocation, 88);
118+
EXPECT_GL_NO_ERROR();
119+
drawQuad(program, essl1_shaders::PositionAttrib(), 0.5f);
120+
glFinish();
121+
EXPECT_GL_ERROR(GL_CONTEXT_LOST);
122+
EXPECT_PIXEL_COLOR_EQ(0, 0, GLColor::transparentBlack); // Should read through client buffer since context should be lost.
123+
}
124+
125+
// Tests that infinite loops based on user-supplied values in vertex shaders hang instead of progress.
126+
// Otherwise optimizer would be able to assume something about the domain of the user-supplied value.
127+
// Explicit value break variant.
128+
TEST_P(TimeoutDrawTest, DynamicInfiniteLoop2VS)
129+
{
130+
constexpr char kVS[] = R"(precision highp float;
131+
attribute vec4 a_position;
132+
uniform int f;
133+
void main()
134+
{
135+
for (;;) { if (f <= 1) break; }
136+
gl_Position = a_position;
137+
})";
138+
ANGLE_GL_PROGRAM(program, kVS, essl1_shaders::fs::Red());
139+
glUseProgram(program);
140+
GLint uniformLocation = glGetUniformLocation(program, "f");
141+
glUniform1i(uniformLocation, 66);
142+
EXPECT_GL_NO_ERROR();
143+
drawQuad(program, essl1_shaders::PositionAttrib(), 0.5f);
144+
glFinish();
145+
EXPECT_GL_ERROR(GL_CONTEXT_LOST);
146+
EXPECT_PIXEL_COLOR_EQ(0, 0, GLColor::transparentBlack); // Should read through client buffer since context should be lost.
147+
}
148+
149+
// Tests that infinite loops based on user-supplied values in fragment shaders hang instead of progress.
150+
// Otherwise optimizer would be able to assume something about the domain of the user-supplied value.
151+
// Explicit value break variant.
152+
TEST_P(TimeoutDrawTest, DynamicInfiniteLoop2FS)
153+
{
154+
constexpr char kFS[] = R"(precision mediump float;
155+
uniform float f;
156+
void main()
157+
{
158+
for (;;) { if (f < 0.1) break; }
159+
gl_FragColor = vec4(1, 0, f, 1);
160+
})";
161+
ANGLE_GL_PROGRAM(program, essl1_shaders::vs::Simple(), kFS);
162+
glUseProgram(program);
163+
GLint uniformLocation = glGetUniformLocation(program, "f");
164+
glUniform1f(uniformLocation, .5f);
165+
EXPECT_GL_NO_ERROR();
166+
drawQuad(program, essl1_shaders::PositionAttrib(), 0.5f);
167+
glFinish();
168+
EXPECT_GL_ERROR(GL_CONTEXT_LOST);
169+
EXPECT_PIXEL_COLOR_EQ(0, 0, GLColor::transparentBlack); // Should read through client buffer since context should be lost.
170+
}
171+
172+
}
173+
174+
ANGLE_INSTANTIATE_TEST(TimeoutDrawTest,
175+
WithRobustness(ES2_METAL().enable(Feature::InjectAsmStatementIntoLoopBodies)),
176+
WithRobustness(ES3_METAL().enable(Feature::InjectAsmStatementIntoLoopBodies)));

0 commit comments

Comments
 (0)