Attachment 373248 Details for Bug 199093 – patch

[patch] patch

b-backup.diff (text/plain), 16.89 KB, created by Saam Barati on 2019-07-01 12:29:15 PDT

(hide)

Description:

Filename:

MIME Type:

Creator: Saam Barati

Created: 2019-07-01 12:29:15 PDT

Size: 16.89 KB

patch

obsolete

>Index: Source/WebCore/ChangeLog
>===================================================================
>--- Source/WebCore/ChangeLog	(revision 247013)
>+++ Source/WebCore/ChangeLog	(working copy)
>@@ -1,3 +1,19 @@
>+2019-07-01  Saam Barati  <sbarati@apple.com>
>+
>+        [WHLSL] Import bitwise bool tests
>+        https://bugs.webkit.org/show_bug.cgi?id=199093
>+
>+        Reviewed by NOBODY (OOPS!).
>+
>+        Add standard library functions for:
>+        - bool bit ops
>+        - converting from bool to number
>+        - converting from number to bool
>+
>+        Test: webgpu/whlsl-bitwise-bool-ops.html
>+
>+        * Modules/webgpu/WHLSL/WHLSLStandardLibrary.txt:
>+
> 2019-07-01  Wenson Hsieh  <wenson_hsieh@apple.com>
> 
>         iOS: REGRESSION(async scroll): Caret doesn't scroll when scrolling textarea
>Index: Source/WebCore/Modules/webgpu/WHLSL/WHLSLStandardLibrary.txt
>===================================================================
>--- Source/WebCore/Modules/webgpu/WHLSL/WHLSLStandardLibrary.txt	(revision 246889)
>+++ Source/WebCore/Modules/webgpu/WHLSL/WHLSLStandardLibrary.txt	(working copy)
>@@ -744,6 +744,12 @@ operator uint2(uint x, uint y) {
>     return result;
> }
> 
>+native bool operator==(uchar, uchar);
>+native bool operator==(ushort, ushort);
>+native bool operator==(char, char);
>+native bool operator==(short, short);
>+native bool operator==(half, half);
>+
> native int operator+(int, int);
> native int operator-(int, int);
> native int operator*(int, int);
>@@ -1172,4 +1178,52 @@ native void GetDimensions(Texture2D<floa
> native void GetDimensions(Texture2D<float4>, uint MipLevel, threadgroup uint* Width, threadgroup uint* Height, device uint* NumberOfLevels);
> native void GetDimensions(Texture2D<float4>, uint MipLevel, threadgroup uint* Width, threadgroup uint* Height, threadgroup uint* NumberOfLevels);
> 
>+
>+native bool operator==(bool, bool);
>+native bool operator&&(bool, bool);
>+native bool operator||(bool, bool);
>+bool operator&(bool a, bool b) {
>+    return a && b;
>+}
>+bool operator|(bool a, bool b) {
>+    return a || b;
>+}
>+bool operator^(bool a, bool b) {
>+    if (a)
>+        return !b;
>+    return b;
>+}
>+bool operator~(bool value) {
>+    return !value;
>+}
>+
>+operator bool(uchar x) {
>+    return x != 0;
>+}
>+operator bool(ushort x) {
>+    return x != 0;
>+}
>+operator bool(uint x) {
>+    return x != 0;
>+}
>+operator bool(char x) {
>+    return x != 0;
>+}
>+operator bool(short x) {
>+    return x != 0;
>+}
>+operator bool(int x) {
>+    return x != 0;
>+}
>+operator bool(half x) {
>+    return x != 0;
>+}
>+operator bool(float x) {
>+    return x != 0;
>+}
>+
>+operator int(bool x) {
>+    return x ? 1 : 0;
>+}
>+
> // FIXME: https://bugs.webkit.org/show_bug.cgi?id=192890 Insert the rest of the standard library once the parser is fast enough
>Index: LayoutTests/ChangeLog
>===================================================================
>--- LayoutTests/ChangeLog	(revision 246891)
>+++ LayoutTests/ChangeLog	(working copy)
>@@ -1,3 +1,37 @@
>+2019-07-01  Saam Barati  <sbarati@apple.com>
>+
>+        [WHLSL] Import bitwise bool tests
>+        https://bugs.webkit.org/show_bug.cgi?id=199093
>+
>+        Reviewed by NOBODY (OOPS!).
>+
>+        This patch makes it so that we can mark bools as input and output types in the
>+        WHLSL harness. Since bool is not something WHLSL itself allows as an entrypoint
>+        input/output type (because we don't specify its bit pattern), we convert between
>+        bool and int in the input and output of the function. For now,  we don't support
>+        a buffer of bools for the input type as a simplification, so we don't have to worry
>+        about dynamically converting an int buffer to a bool buffer. We could add this
>+        in the future if we found it helpful, but we don't have a strong reason for supporting
>+        it right now.
>+        
>+        This patch also starts the process of importing the WHLSL test suite by importing bool
>+        bit op tests.
>+
>+        * webgpu/js/whlsl-test-harness.js:
>+        (convertTypeToArrayType):
>+        (whlslArgumentType):
>+        (convertToWHLSLOutputType):
>+        (convertToWHLSLInputType):
>+        (Data):
>+        (Harness.prototype.get isWHLSL):
>+        (Harness.prototype.async.callTypedFunction):
>+        (Harness.prototype._setUpArguments):
>+        (callVoidFunction):
>+        * webgpu/whlsl-bitwise-bool-ops-expected.txt: Added.
>+        * webgpu/whlsl-bitwise-bool-ops.html: Added.
>+        * webgpu/whlsl-test-harness-test-expected.txt:
>+        * webgpu/whlsl-test-harness-test.html:
>+
> 2019-06-27  Saam Barati  <sbarati@apple.com>
> 
>         Unreviewed. Skip WebGPU tests on High Sierra since WebGPU is disabled.
>Index: LayoutTests/webgpu/whlsl-bitwise-bool-ops-expected.txt
>===================================================================
>--- LayoutTests/webgpu/whlsl-bitwise-bool-ops-expected.txt	(nonexistent)
>+++ LayoutTests/webgpu/whlsl-bitwise-bool-ops-expected.txt	(working copy)
>@@ -0,0 +1,6 @@
>+
>+PASS Bool bit and 
>+PASS Bool bit or 
>+PASS Bool bit or 
>+PASS Bool bit not 
>+
>Index: LayoutTests/webgpu/whlsl-bitwise-bool-ops.html
>===================================================================
>--- LayoutTests/webgpu/whlsl-bitwise-bool-ops.html	(nonexistent)
>+++ LayoutTests/webgpu/whlsl-bitwise-bool-ops.html	(working copy)
>@@ -0,0 +1,175 @@
>+<!DOCTYPE html>
>+<html>
>+<meta charset=utf-8>
>+<meta name="timeout" content="long">
>+<title>Test the WHLSL test harness.</title>
>+<script src="js/whlsl-test-harness.js"></script>
>+<script src="../resources/testharness.js"></script>
>+<script src="../resources/testharnessreport.js"></script>
>+<script>
>+const epsilon = 0.0001;
>+
>+const whlslTests = {};
>+
>+whlslTests.boolBitAnd = () => {
>+    const source = `
>+        bool foo(bool a, bool b)
>+        {
>+            return a & b;
>+        }
>+    `;
>+
>+    webGPUPromiseTest(async () => {
>+        try {
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true), makeBool(true)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false), makeBool(false)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true), makeBool(false)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false), makeBool(true)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+        } catch(e) {
>+            if (!(e instanceof WebGPUUnsupportedError))
>+                throw e;
>+        }
>+    }, "Bool bit and");
>+};
>+
>+whlslTests.boolBitOr = () => {
>+    const source = `
>+        bool foo(bool a, bool b)
>+        {
>+            return a | b;
>+        }
>+    `;
>+
>+    webGPUPromiseTest(async () => {
>+        try {
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true), makeBool(true)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false), makeBool(false)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true), makeBool(false)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false), makeBool(true)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+        } catch(e) {
>+            if (!(e instanceof WebGPUUnsupportedError))
>+                throw e;
>+        }
>+    }, "Bool bit or");
>+};
>+
>+whlslTests.boolBitXor = () => {
>+    const source = `
>+        bool foo(bool a, bool b)
>+        {
>+            return a ^ b;
>+        }
>+    `;
>+
>+    webGPUPromiseTest(async () => {
>+        try {
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true), makeBool(true)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false), makeBool(false)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true), makeBool(false)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false), makeBool(true)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+        } catch(e) {
>+            if (!(e instanceof WebGPUUnsupportedError))
>+                throw e;
>+        }
>+    }, "Bool bit or");
>+};
>+
>+whlslTests.boolBitNot = () => {
>+    const source = `
>+        bool foo(bool a)
>+        {
>+            return ~a;
>+        }
>+    `;
>+
>+    webGPUPromiseTest(async () => {
>+        try {
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(true)]);
>+                assert_equals(result, false, "Test returned expected value.");
>+            }
>+
>+            {
>+                let result = await callBoolFunction(source, "foo", [makeBool(false)]);
>+                assert_equals(result, true, "Test returned expected value.");
>+            }
>+        } catch(e) {
>+            if (!(e instanceof WebGPUUnsupportedError))
>+                throw e;
>+        }
>+    }, "Bool bit not");
>+};
>+
>+function runTests(obj) {
>+    window.addEventListener("load", () => {
>+        try {
>+            for (const name in obj) {
>+                if (!name.startsWith("_"))
>+                    obj[name]();
>+            }
>+        } catch (e) {
>+            if (window.testRunner)
>+                testRunner.notifyDone();
>+            
>+            throw e;
>+        }
>+    });
>+}
>+
>+runTests(whlslTests);
>+
>+const webGPUPromiseTest = (testFunc, msg) => {
>+    promise_test(async () => { 
>+        return testFunc().catch(e => {
>+        if (!(e instanceof WebGPUUnsupportedError))
>+            throw e;
>+        });
>+    }, msg);
>+}
>+</script>
>+</html>
>Index: LayoutTests/webgpu/whlsl-test-harness-test-expected.txt
>===================================================================
>--- LayoutTests/webgpu/whlsl-test-harness-test-expected.txt	(revision 246889)
>+++ LayoutTests/webgpu/whlsl-test-harness-test-expected.txt	(working copy)
>@@ -1,12 +1,15 @@
> 
>+PASS Return a literal of type bool. 
> PASS Return an expected float4 value. 
> PASS Return an expected int value. 
> PASS Return an expected uint value. 
> PASS Return an expected float value. 
>+PASS Upload and return a bool value. 
> PASS Return an expected float4 value. 
> PASS Return an expected int value. 
> PASS Return an expected uint value. 
> PASS Return an expected float value. 
>+PASS Upload many bool values and return a calculated result. 
> PASS Return an expected float4 value. 
> PASS Return an expected int value. 
> PASS Return an expected uint value. 
>Index: LayoutTests/webgpu/whlsl-test-harness-test.html
>===================================================================
>--- LayoutTests/webgpu/whlsl-test-harness-test.html	(revision 246889)
>+++ LayoutTests/webgpu/whlsl-test-harness-test.html	(working copy)
>@@ -58,16 +58,11 @@ whlslTests.manyArguments = () => {
> 
> whlslTests.buffersWithOneValue = () => {
>     const body = `return in0[0];`
>-    checkBools("Access and return a single bool through a bool[].", body, [[true]]);
>     checkFloat4s(body, [[[0, 1, 2, 3]]]);
>     checkNumericScalars(body, [[42]], 42);
> };
> 
> whlslTests.multipleBufferArguments = () => {
>-    checkBools("Access multiple bools through various buffers and return a bool.", 
>-        "return in0[0] & in0[1] & in0[2] & in1 & in2[0];", 
>-        [[true, true, true], true, [true]]);
>-
>     let body = `
>     float x = in0[0].x + in0[1].x + in0[2].x + in1.x + in2[0].x;
>     float y = in0[0].y + in0[1].y + in0[2].y + in1.y + in2[0].y;
>@@ -168,9 +163,6 @@ const checkNumericScalars = (body, argVa
> };
> 
> const checkBools = (msg = "Return an expected bool value.", body, argValues = [], expected = true) => {
>-    // FIXME (https://webkit.org/b/199093): Bool[] functions don't compile, so no-op for now.
>-    return;
>-
>     const [src, name, values] = appendScalarFunctionToSource("", "bool", body, argValues); 
> 
>     webGPUPromiseTest(async () => {
>@@ -231,4 +223,4 @@ const webGPUPromiseTest = (testFunc, msg
>     }, msg);
> }
> </script>
>-</html>
>\ No newline at end of file
>+</html>
>Index: LayoutTests/webgpu/js/whlsl-test-harness.js
>===================================================================
>--- LayoutTests/webgpu/js/whlsl-test-harness.js	(revision 246889)
>+++ LayoutTests/webgpu/js/whlsl-test-harness.js	(working copy)
>@@ -22,10 +22,12 @@ function isVectorType(type)
>     }
> }
> 
>-function convertTypeToArrayType(type)
>+function convertTypeToArrayType(isWHLSL, type)
> {
>     switch(type) {
>         case Types.BOOL:
>+            if (isWHLSL)
>+                return Int32Array;
>             return Uint8Array;
>         case Types.INT:
>             return Int32Array;
>@@ -61,6 +63,27 @@ function convertTypeToWHLSLType(type)
>     }
> }
> 
>+function whlslArgumentType(type)
>+{
>+    if (type === Types.BOOL)
>+        return "int";
>+    return convertTypeToWHLSLType(type);
>+}
>+
>+function convertToWHLSLOutputType(code, type)
>+{
>+    if (type !== Types.BOOL)
>+        return code;
>+    return `int(${code})`;
>+}
>+
>+function convertToWHLSLInputType(code, type)
>+{
>+    if (type !== Types.BOOL)
>+        return code;
>+    return `bool(${code})`;
>+}
>+
> /* Harness Classes */
> 
> class WebGPUUnsupportedError extends Error {
>@@ -91,14 +114,14 @@ class Data {
>         }
> 
>         this._type = type;
>-        this._byteLength = (convertTypeToArrayType(type)).BYTES_PER_ELEMENT * values.length;
>+        this._byteLength = (convertTypeToArrayType(harness.isWHLSL, type)).BYTES_PER_ELEMENT * values.length;
> 
>         const [buffer, arrayBuffer] = harness.device.createBufferMapped({
>             size: this._byteLength,
>             usage: GPUBufferUsage.STORAGE | GPUBufferUsage.MAP_READ
>         });
> 
>-        const typedArray = new (convertTypeToArrayType(type))(arrayBuffer);
>+        const typedArray = new (convertTypeToArrayType(harness.isWHLSL, type))(arrayBuffer);
>         typedArray.set(values);
>         buffer.unmap();
> 
>@@ -157,6 +180,11 @@ using namespace metal;
>         `;
>     }
> 
>+    get isWHLSL()
>+    {
>+        return this._isWHLSL;
>+    }
>+
>     /**
>      * Return the return value of a WHLSL function.
>      * @param {Types} type - The return type of the WHLSL function.
>@@ -196,12 +224,14 @@ using namespace metal;
> 
>         let entryPointCode;
>         if (this._isWHLSL) {
>-            argsDeclarations.unshift(`device ${convertTypeToWHLSLType(type)}[] result : register(u0)`);
>+            argsDeclarations.unshift(`device ${whlslArgumentType(type)}[] result : register(u0)`);
>+            let callCode = `${name}(${functionCallArgs.join(", ")})`;
>+            callCode = convertToWHLSLOutputType(callCode, type);
>             entryPointCode = `
> [numthreads(1, 1, 1)]
> compute void _compute_main(${argsDeclarations.join(", ")})
> {
>-    result[0] = ${name}(${functionCallArgs.join(", ")});
>+    result[0] = ${callCode};
> }
> `;
>         } else {
>@@ -225,7 +255,7 @@ kernel void _compute_main(device _comput
>         } catch {
>             throw new Error("Harness error: Unable to read results!");
>         }
>-        const array = new (convertTypeToArrayType(type))(result);
>+        const array = new (convertTypeToArrayType(this._isWHLSL, type))(result);
>         this._resultBuffer.unmap();
> 
>         return array;
>@@ -301,8 +331,8 @@ kernel void _compute_main(device _comput
>         for (let i = 1; i <= args.length; ++i) {
>             const arg = args[i - 1];
>             if (this._isWHLSL) {
>-                argsDeclarations.push(`device ${convertTypeToWHLSLType(arg.type)}[] arg${i} : register(u${i})`);
>-                functionCallArgs.push(`arg${i}` + (arg.isBuffer ? "" : "[0]"));
>+                argsDeclarations.push(`device ${whlslArgumentType(arg.type)}[] arg${i} : register(u${i})`);
>+                functionCallArgs.push(convertToWHLSLInputType(`arg${i}` + (arg.isBuffer ? "" : "[0]"), arg.type));
>             } else {
>                 argsDeclarations.push(`device ${convertTypeToWHLSLType(arg.type)}* arg${i} [[id(${i})]];`);
>                 functionCallArgs.push((arg.isBuffer ? "" : "*") + `args.arg${i}`);
>@@ -456,4 +486,4 @@ async function callFloat4Function(functi
> function callVoidFunction(functions, name, args)
> {
>     harness.callVoidFunction(functions, name, args);
>-}
>\ No newline at end of file
>+}

Actions: View | Formatted Diff | Diff

Attachments on bug 199093: 373056 | 373248