CopyTextureForBrowser: Support color format conversion

This CL enables blit from RGBA8Unorm soruce texture to dst texture that
|CopyImageBitmapToTexture| supported dst format.

BUG=dawn:465

Change-Id: I99846cf8dc37bc89e0c168a3d86193bb3a0c0ebb
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/44020
Commit-Queue: Shaobo Yan <shaobo.yan@intel.com>
Reviewed-by: Corentin Wallez <cwallez@chromium.org>
Reviewed-by: Austin Eng <enga@chromium.org>

src/dawn_native/CommandValidation.cpp

@@ -392,7 +392,7 @@ namespace dawn_native {
         return {};
     }
 
-    MaybeError ValidateTextureToTextureCopyRestrictions(const ImageCopyTexture& src,
+    MaybeError ValidateTextureToTextureCopyCommonRestrictions(const ImageCopyTexture& src,
                                                               const ImageCopyTexture& dst,
                                                               const Extent3D& copySize) {
         const uint32_t srcSamples = src.texture->GetSampleCount();
@@ -403,11 +403,6 @@ namespace dawn_native {
                 "Source and destination textures must have matching sample counts.");
         }
 
-        if (src.texture->GetFormat().format != dst.texture->GetFormat().format) {
-            // Metal requires texture-to-texture copies be the same format
-            return DAWN_VALIDATION_ERROR("Source and destination texture formats must match.");
-        }
-
         // Metal cannot select a single aspect for texture-to-texture copies.
         const Format& format = src.texture->GetFormat();
         if (SelectFormatAspects(format, src.aspect) != format.aspects) {
@@ -432,6 +427,34 @@ namespace dawn_native {
         return {};
     }
 
+    MaybeError ValidateTextureToTextureCopyRestrictions(const ImageCopyTexture& src,
+                                                        const ImageCopyTexture& dst,
+                                                        const Extent3D& copySize) {
+        if (src.texture->GetFormat().format != dst.texture->GetFormat().format) {
+            // Metal requires texture-to-texture copies be the same format
+            return DAWN_VALIDATION_ERROR("Source and destination texture formats must match.");
+        }
+
+        return ValidateTextureToTextureCopyCommonRestrictions(src, dst, copySize);
+    }
+
+    // CopyTextureForBrowser could handle color conversion during the copy and it
+    // requires the source must be sampleable and the destination must be writable
+    // using a render pass
+    MaybeError ValidateCopyTextureForBrowserRestrictions(const ImageCopyTexture& src,
+                                                         const ImageCopyTexture& dst,
+                                                         const Extent3D& copySize) {
+        if (!(src.texture->GetUsage() & wgpu::TextureUsage::Sampled)) {
+            return DAWN_VALIDATION_ERROR("Source texture must have sampled usage");
+        }
+
+        if (!(dst.texture->GetUsage() & wgpu::TextureUsage::OutputAttachment)) {
+            return DAWN_VALIDATION_ERROR("Dest texture must have outputAttachment usage");
+        }
+
+        return ValidateTextureToTextureCopyCommonRestrictions(src, dst, copySize);
+    }
+
     MaybeError ValidateCanUseAs(const TextureBase* texture, wgpu::TextureUsage usage) {
         ASSERT(wgpu::HasZeroOrOneBits(usage));
         if (!(texture->GetUsage() & usage)) {

src/dawn_native/CommandValidation.h

@@ -75,6 +75,10 @@ namespace dawn_native {
                                                         const ImageCopyTexture& dst,
                                                         const Extent3D& copySize);
 
+    MaybeError ValidateCopyTextureForBrowserRestrictions(const ImageCopyTexture& src,
+                                                         const ImageCopyTexture& dst,
+                                                         const Extent3D& copySize);
+
     MaybeError ValidateCanUseAs(const TextureBase* texture, wgpu::TextureUsage usage);
 
     MaybeError ValidateCanUseAs(const BufferBase* buffer, wgpu::BufferUsage usage);

src/dawn_native/CopyTextureForBrowserHelper.cpp

@@ -32,8 +32,8 @@
 
 namespace dawn_native {
     namespace {
-        // TODO(shaobo.yan@intel.com) : Support premultiplay-alpha, flipY.
+        // TODO(shaobo.yan@intel.com) : Support premultiplay-alpha
-        static const char sCopyTextureForBrowserVertex[] = R"(
+        static const std::string sCopyTextureForBrowserVertex = R"(
             [[block]] struct Uniforms {
                 u_scale : vec2<f32>;
                 u_offset : vec2<f32>;
@@ -56,25 +56,31 @@ namespace dawn_native {
             }
         )";
 
-        static const char sPassthrough2D4ChannelFrag[] = R"(
+        static const std::string sCopyTextureForBrowserFragment = R"(
             [[binding(1), group(0)]] var mySampler: sampler;
             [[binding(2), group(0)]] var myTexture: texture_2d<f32>;
             [[location(0)]] var<in> v_texcoord : vec2<f32>;
-            [[location(0)]] var<out> rgbaColor : vec4<f32>;
+            [[location(0)]] var<out> outputColor : vec4<f32>;
             [[stage(fragment)]] fn main() -> void {
                 // Clamp the texcoord and discard the out-of-bound pixels.
                 var clampedTexcoord : vec2<f32> =
                     clamp(v_texcoord, vec2<f32>(0.0, 0.0), vec2<f32>(1.0, 1.0));
                 if (all(clampedTexcoord == v_texcoord)) {
-                    rgbaColor = textureSample(myTexture, mySampler, v_texcoord);
+                    var srcColor : vec4<f32> = textureSample(myTexture, mySampler, v_texcoord);
+                    // Swizzling of texture formats when sampling / rendering is handled by the
+                    // hardware so we don't need special logic in this shader. This is covered by tests.
+                    outputColor = srcColor;
                 }
             }
         )";
 
-        // TODO(shaobo.yan@intel.com): Expand supported texture formats
+        // TODO(shaobo.yan@intel.com): Expand copyTextureForBrowser to support any
+        // non-depth, non-stencil, non-compressed texture format pair copy. Now this API
+        // supports CopyImageBitmapToTexture normal format pairs.
         MaybeError ValidateCopyTextureFormatConversion(const wgpu::TextureFormat srcFormat,
                                                        const wgpu::TextureFormat dstFormat) {
             switch (srcFormat) {
+                case wgpu::TextureFormat::BGRA8Unorm:
                 case wgpu::TextureFormat::RGBA8Unorm:
                     break;
                 default:
@@ -84,6 +90,12 @@ namespace dawn_native {
 
             switch (dstFormat) {
                 case wgpu::TextureFormat::RGBA8Unorm:
+                case wgpu::TextureFormat::BGRA8Unorm:
+                case wgpu::TextureFormat::RGBA32Float:
+                case wgpu::TextureFormat::RG8Unorm:
+                case wgpu::TextureFormat::RGBA16Float:
+                case wgpu::TextureFormat::RG16Float:
+                case wgpu::TextureFormat::RGB10A2Unorm:
                     break;
                 default:
                     return DAWN_VALIDATION_ERROR(
@@ -103,15 +115,26 @@ namespace dawn_native {
             return {};
         }
 
-        RenderPipelineBase* GetOrCreateCopyTextureForBrowserPipeline(DeviceBase* device) {
+        RenderPipelineBase* GetCachedPipeline(InternalPipelineStore* store,
+                                              wgpu::TextureFormat dstFormat) {
+            auto pipeline = store->copyTextureForBrowserPipelines.find(dstFormat);
+            if (pipeline != store->copyTextureForBrowserPipelines.end()) {
+                return pipeline->second.Get();
+            }
+            return nullptr;
+        }
+
+        RenderPipelineBase* GetOrCreateCopyTextureForBrowserPipeline(
+            DeviceBase* device,
+            wgpu::TextureFormat dstFormat) {
             InternalPipelineStore* store = device->GetInternalPipelineStore();
 
-            if (store->copyTextureForBrowserPipeline == nullptr) {
+            if (GetCachedPipeline(store, dstFormat) == nullptr) {
                 // Create vertex shader module if not cached before.
                 if (store->copyTextureForBrowserVS == nullptr) {
                     ShaderModuleDescriptor descriptor;
                     ShaderModuleWGSLDescriptor wgslDesc;
-                    wgslDesc.source = sCopyTextureForBrowserVertex;
+                    wgslDesc.source = sCopyTextureForBrowserVertex.c_str();
                     descriptor.nextInChain = reinterpret_cast<ChainedStruct*>(&wgslDesc);
 
                     store->copyTextureForBrowserVS =
@@ -124,7 +147,7 @@ namespace dawn_native {
                 if (store->copyTextureForBrowserFS == nullptr) {
                     ShaderModuleDescriptor descriptor;
                     ShaderModuleWGSLDescriptor wgslDesc;
-                    wgslDesc.source = sPassthrough2D4ChannelFrag;
+                    wgslDesc.source = sCopyTextureForBrowserFragment.c_str();
                     descriptor.nextInChain = reinterpret_cast<ChainedStruct*>(&wgslDesc);
                     store->copyTextureForBrowserFS =
                         AcquireRef(device->CreateShaderModule(&descriptor));
@@ -144,7 +167,7 @@ namespace dawn_native {
 
                 // Prepare color state.
                 ColorTargetState target = {};
-                target.format = wgpu::TextureFormat::RGBA8Unorm;
+                target.format = dstFormat;
 
                 // Create RenderPipeline.
                 RenderPipelineDescriptor2 renderPipelineDesc = {};
@@ -160,11 +183,11 @@ namespace dawn_native {
                 fragment.targetCount = 1;
                 fragment.targets = &target;
 
-                store->copyTextureForBrowserPipeline =
+                store->copyTextureForBrowserPipelines.insert(
-                    AcquireRef(device->CreateRenderPipeline2(&renderPipelineDesc));
+                    {dstFormat, AcquireRef(device->CreateRenderPipeline2(&renderPipelineDesc))});
             }
 
-            return store->copyTextureForBrowserPipeline.Get();
+            return GetCachedPipeline(store, dstFormat);
         }
 
     }  // anonymous namespace
@@ -180,7 +203,7 @@ namespace dawn_native {
         DAWN_TRY(ValidateImageCopyTexture(device, *source, *copySize));
         DAWN_TRY(ValidateImageCopyTexture(device, *destination, *copySize));
 
-        DAWN_TRY(ValidateTextureToTextureCopyRestrictions(*source, *destination, *copySize));
+        DAWN_TRY(ValidateCopyTextureForBrowserRestrictions(*source, *destination, *copySize));
 
         DAWN_TRY(ValidateTextureCopyRange(*source, *copySize));
         DAWN_TRY(ValidateTextureCopyRange(*destination, *copySize));
@@ -214,7 +237,9 @@ namespace dawn_native {
                                        const CopyTextureForBrowserOptions* options) {
         // TODO(shaobo.yan@intel.com): In D3D12 and Vulkan, compatible texture format can directly
         // copy to each other. This can be a potential fast path.
-        RenderPipelineBase* pipeline = GetOrCreateCopyTextureForBrowserPipeline(device);
+
+        RenderPipelineBase* pipeline = GetOrCreateCopyTextureForBrowserPipeline(
+            device, destination->texture->GetFormat().format);
 
         // Prepare bind group layout.
         Ref<BindGroupLayoutBase> layout = AcquireRef(pipeline->GetBindGroupLayout(0));
@@ -232,7 +257,7 @@ namespace dawn_native {
             0.0, 0.0   // offset
         };
 
-        // Handle flipY.
+        // Handle flipY
         if (options && options->flipY) {
             uniformData[1] *= -1.0;
             uniformData[3] += 1.0;
@@ -282,6 +307,7 @@ namespace dawn_native {
 
         // Prepare render pass color attachment descriptor.
         RenderPassColorAttachmentDescriptor colorAttachmentDesc;
+
         colorAttachmentDesc.attachment = dstView.Get();
         colorAttachmentDesc.loadOp = wgpu::LoadOp::Load;
         colorAttachmentDesc.storeOp = wgpu::StoreOp::Store;

src/dawn_native/InternalPipelineStore.h

@@ -25,7 +25,9 @@ namespace dawn_native {
     class ShaderModuleBase;
 
     struct InternalPipelineStore {
-        Ref<RenderPipelineBase> copyTextureForBrowserPipeline;
+        std::unordered_map<wgpu::TextureFormat, Ref<RenderPipelineBase>>
+            copyTextureForBrowserPipelines;
+
         Ref<ShaderModuleBase> copyTextureForBrowserVS;
         Ref<ShaderModuleBase> copyTextureForBrowserFS;
 

src/tests/end2end/CopyTextureForBrowserTests.cpp

@@ -21,12 +21,23 @@
 #include "utils/TextureFormatUtils.h"
 #include "utils/WGPUHelpers.h"
 
+namespace {
+    static constexpr wgpu::TextureFormat kTextureFormat = wgpu::TextureFormat::RGBA8Unorm;
+
+    // Set default texture size to single line texture for color conversion tests.
+    static constexpr uint64_t kDefaultTextureWidth = 10;
+    static constexpr uint64_t kDefaultTextureHeight = 1;
+
+    // Dst texture format copyTextureForBrowser accept
+    static constexpr wgpu::TextureFormat kDstTextureFormat[] = {
+        wgpu::TextureFormat::RGBA8Unorm,  wgpu::TextureFormat::BGRA8Unorm,
+        wgpu::TextureFormat::RGBA32Float, wgpu::TextureFormat::RG8Unorm,
+        wgpu::TextureFormat::RGBA16Float, wgpu::TextureFormat::RG16Float,
+        wgpu::TextureFormat::RGB10A2Unorm};
+}  // anonymous namespace
+
 class CopyTextureForBrowserTests : public DawnTest {
   protected:
-    static constexpr wgpu::TextureFormat kTextureFormat = wgpu::TextureFormat::RGBA8Unorm;
-    static constexpr uint64_t kDefaultTextureWidth = 4;
-    static constexpr uint64_t kDefaultTextureHeight = 4;
-
     struct TextureSpec {
         wgpu::Origin3D copyOrigin = {};
         wgpu::Extent3D textureSize = {kDefaultTextureWidth, kDefaultTextureHeight};
@@ -34,6 +45,33 @@ class CopyTextureForBrowserTests : public DawnTest {
         wgpu::TextureFormat format = kTextureFormat;
     };
 
+    // This fixed source texture data is for color conversion tests.
+    // The source data can fill a texture in default width and height.
+    static std::vector<RGBA8> GetFixedSourceTextureData() {
+        std::vector<RGBA8> sourceTextureData{
+            // Take RGBA8Unorm as example:
+            // R channel has different values
+            RGBA8(0, 255, 255, 255),    // r = 0.0
+            RGBA8(102, 255, 255, 255),  // r = 0.4
+            RGBA8(153, 255, 255, 255),  // r = 0.6
+
+            // G channel has different values
+            RGBA8(255, 0, 255, 255),    // g = 0.0
+            RGBA8(255, 102, 255, 255),  // g = 0.4
+            RGBA8(255, 153, 255, 255),  // g = 0.6
+
+            // B channel has different values
+            RGBA8(255, 255, 0, 255),    // b = 0.0
+            RGBA8(255, 255, 102, 255),  // b = 0.4
+            RGBA8(255, 255, 153, 255),  // b = 0.6
+
+            // A channel set to 0
+            RGBA8(255, 255, 255, 0)  // a = 0
+        };
+
+        return sourceTextureData;
+    }
+
     static std::vector<RGBA8> GetSourceTextureData(const utils::TextureDataCopyLayout& layout) {
         std::vector<RGBA8> textureData(layout.texelBlockCount);
         for (uint32_t layer = 0; layer < layout.mipSize.depth; ++layer) {
@@ -44,7 +82,7 @@ class CopyTextureForBrowserTests : public DawnTest {
                     textureData[sliceOffset + rowOffset + x] =
                         RGBA8(static_cast<uint8_t>((x + layer * x) % 256),
                               static_cast<uint8_t>((y + layer * y) % 256),
-                              static_cast<uint8_t>(x / 256), static_cast<uint8_t>(y / 256));
+                              static_cast<uint8_t>(x % 256), static_cast<uint8_t>(x % 256));
                 }
             }
         }
@@ -59,6 +97,7 @@ class CopyTextureForBrowserTests : public DawnTest {
 
         uint32_t uniformBufferData[] = {
             0,  // copy have flipY option
+            4,  // channelCount
         };
 
         wgpu::BufferDescriptor uniformBufferDesc = {};
@@ -74,6 +113,7 @@ class CopyTextureForBrowserTests : public DawnTest {
         wgpu::ShaderModule csModule = utils::CreateShaderModuleFromWGSL(device, R"(
             [[block]] struct Uniforms {
                 dstTextureFlipY : u32;
+                channelCount : u32;
             };
             [[block]] struct OutputBuf {
                 result : array<u32>;
@@ -83,10 +123,13 @@ class CopyTextureForBrowserTests : public DawnTest {
             [[group(0), binding(2)]] var<storage> output : [[access(read_write)]] OutputBuf;
             [[group(0), binding(3)]] var<uniform> uniforms : Uniforms;
             [[builtin(global_invocation_id)]] var<in> GlobalInvocationID : vec3<u32>;
-            [[stage(compute), workgroup_size(1, 1, 1)]]
+            fn aboutEqual(value : f32, expect : f32) -> bool {
-            fn main() -> void {
+                // The value diff should be smaller than the hard coded tolerance.
+                return abs(value - expect) < 0.001;
+            }
+            [[stage(compute), workgroup_size(1, 1, 1)]] fn main() -> void {
                 // Current CopyTextureForBrowser only support full copy now.
-                // TODO(dawn:465): Refactor this after CopyTextureForBrowser
+                // TODO(crbug.com/dawn/465): Refactor this after CopyTextureForBrowser
                 // support sub-rect copy.
                 var size : vec2<i32> = textureDimensions(src);
                 var dstTexCoord : vec2<i32> = vec2<i32>(GlobalInvocationID.xy);
@@ -97,7 +140,21 @@ class CopyTextureForBrowserTests : public DawnTest {
 
                 var srcColor : vec4<f32> = textureLoad(src, srcTexCoord, 0);
                 var dstColor : vec4<f32> = textureLoad(dst, dstTexCoord, 0);
-                var success : bool = all(srcColor == dstColor);
+                var success : bool = true;
+
+                // Not use loop and variable index format to workaround
+                // crbug.com/tint/638.
+                if (uniforms.channelCount == 2u) { // All have rg components.
+                    success = success &&
+                              aboutEqual(dstColor.r, srcColor.r) &&
+                              aboutEqual(dstColor.g, srcColor.g);
+                } else {
+                    success = success &&
+                              aboutEqual(dstColor.r, srcColor.r) &&
+                              aboutEqual(dstColor.g, srcColor.g) &&
+                              aboutEqual(dstColor.b, srcColor.b) &&
+                              aboutEqual(dstColor.a, srcColor.a);
+                }
 
                 var outputIndex : u32 = GlobalInvocationID.y * u32(size.x) + GlobalInvocationID.x;
                 if (success) {
@@ -114,11 +171,31 @@ class CopyTextureForBrowserTests : public DawnTest {
 
         return device.CreateComputePipeline(&csDesc);
     }
+    static uint32_t GetTextureFormatComponentCount(wgpu::TextureFormat format) {
+        switch (format) {
+            case wgpu::TextureFormat::RGBA8Unorm:
+            case wgpu::TextureFormat::BGRA8Unorm:
+            case wgpu::TextureFormat::RGB10A2Unorm:
+            case wgpu::TextureFormat::RGBA16Float:
+            case wgpu::TextureFormat::RGBA32Float:
+                return 4;
+            case wgpu::TextureFormat::RG8Unorm:
+            case wgpu::TextureFormat::RG16Float:
+                return 2;
+            default:
+                UNREACHABLE();
+        }
+    }
+
+    void DoColorConversionTest(const TextureSpec& srcSpec, const TextureSpec& dstSpec) {
+        DoTest(srcSpec, dstSpec, {kDefaultTextureWidth, kDefaultTextureHeight}, {}, true);
+    }
 
     void DoTest(const TextureSpec& srcSpec,
                 const TextureSpec& dstSpec,
                 const wgpu::Extent3D& copySize = {kDefaultTextureWidth, kDefaultTextureHeight},
-                const wgpu::CopyTextureForBrowserOptions options = {}) {
+                const wgpu::CopyTextureForBrowserOptions options = {},
+                bool useFixedTestValue = false) {
         wgpu::TextureDescriptor srcDescriptor;
         srcDescriptor.size = srcSpec.textureSize;
         srcDescriptor.format = srcSpec.format;
@@ -133,7 +210,7 @@ class CopyTextureForBrowserTests : public DawnTest {
         dstDescriptor.format = dstSpec.format;
         dstDescriptor.mipLevelCount = dstSpec.level + 1;
         dstDescriptor.usage = wgpu::TextureUsage::CopyDst | wgpu::TextureUsage::Sampled |
-                              wgpu::TextureUsage::OutputAttachment;
+                              wgpu::TextureUsage::OutputAttachment | wgpu::TextureUsage::CopySrc;
         dstTexture = device.CreateTexture(&dstDescriptor);
 
         wgpu::CommandEncoder encoder = device.CreateCommandEncoder();
@@ -144,7 +221,8 @@ class CopyTextureForBrowserTests : public DawnTest {
                 {srcSpec.textureSize.width, srcSpec.textureSize.height, copySize.depth},
                 srcSpec.level);
 
-        const std::vector<RGBA8> textureArrayCopyData = GetSourceTextureData(copyLayout);
+        const std::vector<RGBA8> textureArrayCopyData =
+            useFixedTestValue ? GetFixedSourceTextureData() : GetSourceTextureData(copyLayout);
         wgpu::ImageCopyTexture imageCopyTexture =
             utils::CreateImageCopyTexture(srcTexture, srcSpec.level, {0, 0, srcSpec.copyOrigin.z});
 
@@ -172,7 +250,7 @@ class CopyTextureForBrowserTests : public DawnTest {
         // Update uniform buffer based on test config
         uint32_t uniformBufferData[] = {
             options.flipY,                                    // copy have flipY option
-        };
+            GetTextureFormatComponentCount(dstSpec.format)};  // channelCount
 
         device.GetQueue().WriteBuffer(uniformBuffer, 0, uniformBufferData,
                                       sizeof(uniformBufferData));
@@ -310,6 +388,45 @@ TEST_P(CopyTextureForBrowserTests, VerifyFlipYInSlimTexture) {
     DoTest(textureSpec, textureSpec, {kWidth, kHeight}, options);
 }
 
+// Verify |CopyTextureForBrowser| doing color conversion correctly when
+// the source texture is RGBA8Unorm format.
+TEST_P(CopyTextureForBrowserTests, FromRGBA8UnormCopy) {
+    // Tests skip due to crbug.com/dawn/592.
+    DAWN_SKIP_TEST_IF(IsD3D12() && IsBackendValidationEnabled());
+    // Skip OpenGLES backend because it fails on using RGBA8Unorm as
+    // source texture format.
+    DAWN_SKIP_TEST_IF(IsOpenGLES());
+
+    for (wgpu::TextureFormat dstFormat : kDstTextureFormat) {
+        TextureSpec srcTextureSpec = {};  // default format is RGBA8Unorm
+
+        TextureSpec dstTextureSpec;
+        dstTextureSpec.format = dstFormat;
+
+        DoColorConversionTest(srcTextureSpec, dstTextureSpec);
+    }
+}
+
+// Verify |CopyTextureForBrowser| doing color conversion correctly when
+// the source texture is BGRAUnorm format.
+TEST_P(CopyTextureForBrowserTests, FromBGRA8UnormCopy) {
+    // Tests skip due to crbug.com/dawn/592.
+    DAWN_SKIP_TEST_IF(IsD3D12() && IsBackendValidationEnabled());
+    // Skip OpenGLES backend because it fails on using BGRA8Unorm as
+    // source texture format.
+    DAWN_SKIP_TEST_IF(IsOpenGLES());
+
+    for (wgpu::TextureFormat dstFormat : kDstTextureFormat) {
+        TextureSpec srcTextureSpec;
+        srcTextureSpec.format = wgpu::TextureFormat::BGRA8Unorm;
+
+        TextureSpec dstTextureSpec;
+        dstTextureSpec.format = dstFormat;
+
+        DoColorConversionTest(srcTextureSpec, dstTextureSpec);
+    }
+}
+
 DAWN_INSTANTIATE_TEST(CopyTextureForBrowserTests,
                       D3D12Backend(),
                       MetalBackend(),