From 5e2e2d863e7ac27fb3c9d9e8ab729268ca0171a3 Mon Sep 17 00:00:00 2001 From: Shrek Shao Date: Thu, 28 Oct 2021 00:04:07 +0000 Subject: [PATCH] Pipeline constants: validate if any uninitialized overridable constants exist Now an overridable constant need to be initialized, either via value specified in shader, or via constant entry from pipeline stage. Otherwise it is invalid. Together fix the bool 32bit initialize problem on vulkan and re-enable the overridable constants shader test, and use assigning to _ instead of ignore() in tests. Bug: dawn:1041 Change-Id: I49e7885c8d6134647b09926ceb15234ac21ee35d Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/67560 Commit-Queue: Shrek Shao Reviewed-by: Corentin Wallez --- src/dawn_native/Pipeline.cpp | 33 +++++ src/dawn_native/ShaderModule.cpp | 19 ++- src/dawn_native/ShaderModule.h | 13 +- src/dawn_native/vulkan/UtilsVulkan.cpp | 8 +- src/dawn_native/vulkan/UtilsVulkan.h | 3 +- src/tests/end2end/ShaderTests.cpp | 26 +--- .../OverridableConstantsValidationTests.cpp | 113 ++++++++++++++---- .../validation/UnsafeAPIValidationTests.cpp | 4 +- 8 files changed, 159 insertions(+), 60 deletions(-) diff --git a/src/dawn_native/Pipeline.cpp b/src/dawn_native/Pipeline.cpp index 96258bc63c..ab232c3566 100644 --- a/src/dawn_native/Pipeline.cpp +++ b/src/dawn_native/Pipeline.cpp @@ -53,10 +53,43 @@ namespace dawn_native { // Validate if overridable constants exist in shader module // pipelineBase is not yet constructed at this moment so iterate constants from descriptor + size_t numUninitializedConstants = metadata.uninitializedOverridableConstants.size(); + // Keep a initialized constants sets to handle duplicate initialization cases + // Only storing that of uninialized constants is needed + std::unordered_set stageInitializedConstantIdentifiers; for (uint32_t i = 0; i < constantCount; i++) { DAWN_INVALID_IF(metadata.overridableConstants.count(constants[i].key) == 0, "Pipeline overridable constant \"%s\" not found in shader module %s.", constants[i].key, module); + + if (metadata.uninitializedOverridableConstants.count(constants[i].key) > 0 && + stageInitializedConstantIdentifiers.count(constants[i].key) == 0) { + numUninitializedConstants--; + stageInitializedConstantIdentifiers.insert(constants[i].key); + } + } + + // Validate if any overridable constant is left uninitialized + if (DAWN_UNLIKELY(numUninitializedConstants > 0)) { + std::string uninitializedConstantsArray; + bool isFirst = true; + for (std::string identifier : metadata.uninitializedOverridableConstants) { + if (stageInitializedConstantIdentifiers.count(identifier) > 0) { + continue; + } + + if (isFirst) { + isFirst = false; + } else { + uninitializedConstantsArray.append(", "); + } + uninitializedConstantsArray.append(identifier); + } + + return DAWN_FORMAT_VALIDATION_ERROR( + "There are uninitialized pipeline overridable constants in shader module %s, their " + "identifiers:[%s]", + module, uninitializedConstantsArray); } return {}; diff --git a/src/dawn_native/ShaderModule.cpp b/src/dawn_native/ShaderModule.cpp index 7a7bbd7133..058f886e48 100644 --- a/src/dawn_native/ShaderModule.cpp +++ b/src/dawn_native/ShaderModule.cpp @@ -639,12 +639,19 @@ namespace dawn_native { for (auto& c : entryPoint.overridable_constants) { EntryPointMetadata::OverridableConstant constant = { - name2Id.at(c.name), FromTintOverridableConstantType(c.type)}; - metadata->overridableConstants[c.name] = constant; - // TODO(tint:1155) tint needs ways to differentiate whether a pipeline - // constant id is specified explicitly. Now we just store numeric id and - // variable name in the index at the same time - metadata->overridableConstants[std::to_string(constant.id)] = constant; + name2Id.at(c.name), FromTintOverridableConstantType(c.type), + c.is_initialized}; + + std::string identifier = + c.is_numeric_id_specified ? std::to_string(constant.id) : c.name; + metadata->overridableConstants[identifier] = constant; + + if (!c.is_initialized) { + auto it = metadata->uninitializedOverridableConstants.emplace( + std::move(identifier)); + // The insertion should have taken place + ASSERT(it.second); + } } } diff --git a/src/dawn_native/ShaderModule.h b/src/dawn_native/ShaderModule.h index 070d001650..09bede54cd 100644 --- a/src/dawn_native/ShaderModule.h +++ b/src/dawn_native/ShaderModule.h @@ -32,6 +32,7 @@ #include #include #include +#include #include namespace tint { @@ -200,11 +201,21 @@ namespace dawn_native { // Match tint::inspector::OverridableConstant::Type // Bool is defined as a macro on linux X11 and cannot compile enum class Type { Boolean, Float32, Uint32, Int32 } type; + + // If the constant doesn't not have an initializer in the shader + // Then it is required for the pipeline stage to have a constant record to initialize a + // value + bool isInitialized; }; - // Store overridableConstants from tint program + // Map identifier to overridable constant + // Identifier is unique: either the variable name or the numeric ID if specified std::unordered_map overridableConstants; + // Overridable constants that are not initialized in shaders + // They need value initialization from pipeline stage or it is a validation error + std::unordered_set uninitializedOverridableConstants; + bool usesNumWorkgroups = false; }; diff --git a/src/dawn_native/vulkan/UtilsVulkan.cpp b/src/dawn_native/vulkan/UtilsVulkan.cpp index 5d9029073f..87d896d4aa 100644 --- a/src/dawn_native/vulkan/UtilsVulkan.cpp +++ b/src/dawn_native/vulkan/UtilsVulkan.cpp @@ -215,11 +215,11 @@ namespace dawn_native { namespace vulkan { programmableStage.module->GetEntryPoint(programmableStage.entryPoint); for (const auto& pipelineConstant : programmableStage.constants) { - const std::string& name = pipelineConstant.first; + const std::string& identifier = pipelineConstant.first; double value = pipelineConstant.second; - // This is already validated so `name` must exist - const auto& moduleConstant = entryPointMetaData.overridableConstants.at(name); + // This is already validated so `identifier` must exist + const auto& moduleConstant = entryPointMetaData.overridableConstants.at(identifier); specializationMapEntries->push_back( VkSpecializationMapEntry{moduleConstant.id, @@ -230,7 +230,7 @@ namespace dawn_native { namespace vulkan { SpecializationDataEntry entry{}; switch (moduleConstant.type) { case EntryPointMetadata::OverridableConstant::Type::Boolean: - entry.b = static_cast(value); + entry.b = static_cast(value); break; case EntryPointMetadata::OverridableConstant::Type::Float32: entry.f32 = static_cast(value); diff --git a/src/dawn_native/vulkan/UtilsVulkan.h b/src/dawn_native/vulkan/UtilsVulkan.h index 8bd2a84544..c3859172d9 100644 --- a/src/dawn_native/vulkan/UtilsVulkan.h +++ b/src/dawn_native/vulkan/UtilsVulkan.h @@ -114,7 +114,8 @@ namespace dawn_native { namespace vulkan { // Helpers for creating VkSpecializationInfo // The WebGPU overridable constants only support these scalar types union SpecializationDataEntry { - bool b; + // Use int32_t for boolean to initialize the full 32bit + int32_t b; float f32; int32_t i32; uint32_t u32; diff --git a/src/tests/end2end/ShaderTests.cpp b/src/tests/end2end/ShaderTests.cpp index 7c32ad6d4b..aa40fe85f2 100644 --- a/src/tests/end2end/ShaderTests.cpp +++ b/src/tests/end2end/ShaderTests.cpp @@ -393,19 +393,13 @@ fn main([[location(0)]] pos : vec4) -> [[builtin(position)]] vec4 { } // Test overridable constants without numeric identifiers -// TODO(tint:1155): Implicit numeric ID is undetermined in tint -TEST_P(ShaderTests, DISABLED_OverridableConstants) { +TEST_P(ShaderTests, OverridableConstants) { // TODO(dawn:1041): Only Vulkan backend is implemented DAWN_TEST_UNSUPPORTED_IF(!IsVulkan()); - uint32_t const kCount = 15; + uint32_t const kCount = 11; std::vector expected(kCount); std::iota(expected.begin(), expected.end(), 0); - // Test last entry with unspecified default value - expected[kCount - 1] = 0u; - expected[kCount - 2] = 0u; - expected[kCount - 3] = 0u; - expected[kCount - 4] = 0u; wgpu::Buffer buffer = CreateBuffer(kCount); std::string shader = R"( @@ -420,13 +414,9 @@ TEST_P(ShaderTests, DISABLED_OverridableConstants) { [[override]] let c8: u32; // type: uint32 [[override]] let c9: u32 = 0u; // default override [[override]] let c10: u32 = 10u; // default -[[override]] let c11: bool; // default unspecified -[[override]] let c12: f32; // default unspecified -[[override]] let c13: i32; // default unspecified -[[override]] let c14: u32; // default unspecified [[block]] struct Buf { - data : array; + data : array; }; [[group(0), binding(0)]] var buf : Buf; @@ -443,10 +433,6 @@ TEST_P(ShaderTests, DISABLED_OverridableConstants) { buf.data[8] = u32(c8); buf.data[9] = u32(c9); buf.data[10] = u32(c10); - buf.data[11] = u32(c11); - buf.data[12] = u32(c12); - buf.data[13] = u32(c13); - buf.data[14] = u32(c14); })"; std::vector constants; @@ -461,10 +447,6 @@ TEST_P(ShaderTests, DISABLED_OverridableConstants) { constants.push_back({nullptr, "c8", 8}); constants.push_back({nullptr, "c9", 9}); // c10 is not assigned, testing default value - // c11 is not assigned, testing unspecified default value - // c12 is not assigned, testing unspecified default value - // c13 is not assigned, testing unspecified default value - // c14 is not assigned, testing unspecified default value wgpu::ComputePipeline pipeline = CreateComputePipeline(shader, "main", &constants); @@ -520,7 +502,7 @@ TEST_P(ShaderTests, OverridableConstantsNumericIdentifiers) { constants.push_back({nullptr, "1001", 1}); constants.push_back({nullptr, "1", 2}); // c3 is not assigned, testing default value - // c4 is not assigned, testing unspecified default value + constants.push_back({nullptr, "1004", 0}); wgpu::ComputePipeline pipeline = CreateComputePipeline(shader, "main", &constants); diff --git a/src/tests/unittests/validation/OverridableConstantsValidationTests.cpp b/src/tests/unittests/validation/OverridableConstantsValidationTests.cpp index af96c394e9..af43f4cf86 100644 --- a/src/tests/unittests/validation/OverridableConstantsValidationTests.cpp +++ b/src/tests/unittests/validation/OverridableConstantsValidationTests.cpp @@ -18,9 +18,37 @@ class ComputePipelineOverridableConstantsValidationTest : public ValidationTest { protected: - void SetUp() override { - ValidationTest::SetUp(); + void SetUpShadersWithDefaultValueConstants() { + computeModule = utils::CreateShaderModule(device, R"( +[[override]] let c0: bool = true; // type: bool +[[override]] let c1: bool = false; // default override +[[override]] let c2: f32 = 0.0; // type: float32 +[[override]] let c3: f32 = 0.0; // default override +[[override]] let c4: f32 = 4.0; // default +[[override]] let c5: i32 = 0; // type: int32 +[[override]] let c6: i32 = 0; // default override +[[override]] let c7: i32 = 7; // default +[[override]] let c8: u32 = 0u; // type: uint32 +[[override]] let c9: u32 = 0u; // default override +[[override(1000)]] let c10: u32 = 10u; // default +[[stage(compute), workgroup_size(1)]] fn main() { + // make sure the overridable constants are not optimized out + _ = u32(c0); + _ = u32(c1); + _ = u32(c2); + _ = u32(c3); + _ = u32(c4); + _ = u32(c5); + _ = u32(c6); + _ = u32(c7); + _ = u32(c8); + _ = u32(c9); + _ = u32(c10); +})"); + } + + void SetUpShadersWithUninitializedConstants() { computeModule = utils::CreateShaderModule(device, R"( [[override]] let c0: bool; // type: bool [[override]] let c1: bool = false; // default override @@ -34,25 +62,19 @@ class ComputePipelineOverridableConstantsValidationTest : public ValidationTest [[override]] let c9: u32 = 0u; // default override [[override(1000)]] let c10: u32 = 10u; // default -[[block]] struct Buf { - data : array; -}; - -[[group(0), binding(0)]] var buf : Buf; - [[stage(compute), workgroup_size(1)]] fn main() { // make sure the overridable constants are not optimized out - buf.data[0] = u32(c0); - buf.data[1] = u32(c1); - buf.data[2] = u32(c2); - buf.data[3] = u32(c3); - buf.data[4] = u32(c4); - buf.data[5] = u32(c5); - buf.data[6] = u32(c6); - buf.data[7] = u32(c7); - buf.data[8] = u32(c8); - buf.data[9] = u32(c9); - buf.data[10] = u32(c10); + _ = u32(c0); + _ = u32(c1); + _ = u32(c2); + _ = u32(c3); + _ = u32(c4); + _ = u32(c5); + _ = u32(c6); + _ = u32(c7); + _ = u32(c8); + _ = u32(c9); + _ = u32(c10); })"); } @@ -71,6 +93,7 @@ class ComputePipelineOverridableConstantsValidationTest : public ValidationTest // Basic constants lookup tests TEST_F(ComputePipelineOverridableConstantsValidationTest, ConstantsIdentifierLookUp) { + SetUpShadersWithDefaultValueConstants(); { // Valid: no constants specified std::vector constants; @@ -106,10 +129,47 @@ TEST_F(ComputePipelineOverridableConstantsValidationTest, ConstantsIdentifierLoo } } +// Test that it is invalid to leave any constants uninitialized +TEST_F(ComputePipelineOverridableConstantsValidationTest, UninitializedConstants) { + SetUpShadersWithUninitializedConstants(); + { + // Error: uninitialized constants exist + std::vector constants; + ASSERT_DEVICE_ERROR(TestCreatePipeline(constants)); + } + { + // Error: uninitialized constants exist + std::vector constants{ + {nullptr, "c0", false}, + {nullptr, "c2", 1}, + // c5 is missing + {nullptr, "c8", 1}, + }; + ASSERT_DEVICE_ERROR(TestCreatePipeline(constants)); + } + { + // Valid: all constants initialized + std::vector constants{ + {nullptr, "c0", false}, + {nullptr, "c2", 1}, + {nullptr, "c5", 1}, + {nullptr, "c8", 1}, + }; + TestCreatePipeline(constants); + } + { + // Valid: all constants initialized (with duplicate initializations) + std::vector constants{ + {nullptr, "c0", false}, {nullptr, "c2", 1}, {nullptr, "c5", 1}, + {nullptr, "c8", 1}, {nullptr, "c2", 2}, + }; + TestCreatePipeline(constants); + } +} + // Test that only explicitly specified numeric ID can be referenced -// TODO(tint:1155): missing feature in tint to differentiate explicitly specified numeric ID -TEST_F(ComputePipelineOverridableConstantsValidationTest, - DISABLED_ConstantsIdentifierExplicitNumericID) { +TEST_F(ComputePipelineOverridableConstantsValidationTest, ConstantsIdentifierExplicitNumericID) { + SetUpShadersWithDefaultValueConstants(); { // Error: constant numeric id not explicitly specified // But could be impliciltly assigned to one of the constants @@ -137,8 +197,13 @@ TEST_F(ComputePipelineOverridableConstantsValidationTest, } // Test that identifiers are unique -// TODO(tint:1155): missing feature in tint to differentiate explicitly specified numeric ID -TEST_F(ComputePipelineOverridableConstantsValidationTest, DISABLED_ConstantsIdentifierUnique) { +TEST_F(ComputePipelineOverridableConstantsValidationTest, ConstantsIdentifierUnique) { + SetUpShadersWithDefaultValueConstants(); + { + // Valid: constant without numeric id can be referenced with variable name + std::vector constants{{nullptr, "c0", 0}}; + TestCreatePipeline(constants); + } { // Error: constant with numeric id cannot be referenced with variable name std::vector constants{{nullptr, "c10", 0}}; diff --git a/src/tests/unittests/validation/UnsafeAPIValidationTests.cpp b/src/tests/unittests/validation/UnsafeAPIValidationTests.cpp index a7e915a3bd..624305b015 100644 --- a/src/tests/unittests/validation/UnsafeAPIValidationTests.cpp +++ b/src/tests/unittests/validation/UnsafeAPIValidationTests.cpp @@ -129,8 +129,8 @@ TEST_F(UnsafeAPIValidationTest, PipelineOverridableConstants) { [[override(1000)]] let c1: u32; [[stage(compute), workgroup_size(1)]] fn main() { - ignore(c0); - ignore(c1); + _ = c0; + _ = c1; })")); }