diff --git a/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp b/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp index b7628cdbe7..8fd2d60b95 100644 --- a/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp +++ b/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp @@ -132,7 +132,6 @@ MaybeError BindGroupTracker::ApplyBindGroup(BindGroupIndex index) { switch (bindingInfo.bindingType) { case BindingInfoType::Buffer: { BufferBinding binding = group->GetBindingAsBufferBinding(bindingIndex); - ID3D11Buffer* d3d11Buffer = ToBackend(binding.buffer)->GetD3D11Buffer(); auto offset = binding.offset; if (bindingInfo.buffer.hasDynamicOffset) { // Dynamic buffers are packed at the front of BindingIndices. @@ -141,6 +140,9 @@ MaybeError BindGroupTracker::ApplyBindGroup(BindGroupIndex index) { switch (bindingInfo.buffer.type) { case wgpu::BufferBindingType::Uniform: { + ToBackend(binding.buffer)->EnsureConstantBufferIsUpdated(mCommandContext); + ID3D11Buffer* d3d11Buffer = + ToBackend(binding.buffer)->GetD3D11ConstantBuffer(); // https://learn.microsoft.com/en-us/windows/win32/api/d3d11_1/nf-d3d11_1-id3d11devicecontext1-vssetconstantbuffers1 // Offset and size are measured in shader constants, which are 16 bytes // (4*32-bit components). And the offsets and counts must be multiples @@ -175,6 +177,7 @@ MaybeError BindGroupTracker::ApplyBindGroup(BindGroupIndex index) { DAWN_TRY_ASSIGN( d3d11UAV, ToBackend(binding.buffer) ->CreateD3D11UnorderedAccessView1(offset, binding.size)); + ToBackend(binding.buffer)->MarkMutated(); if (bindingInfo.visibility & wgpu::ShaderStage::Fragment) { deviceContext1->OMSetRenderTargetsAndUnorderedAccessViews( D3D11_KEEP_RENDER_TARGETS_AND_DEPTH_STENCIL, nullptr, nullptr, diff --git a/src/dawn/native/d3d11/BufferD3D11.cpp b/src/dawn/native/d3d11/BufferD3D11.cpp index b399870bef..233c195e0d 100644 --- a/src/dawn/native/d3d11/BufferD3D11.cpp +++ b/src/dawn/native/d3d11/BufferD3D11.cpp @@ -33,19 +33,8 @@ namespace dawn::native::d3d11 { namespace { -MaybeError ValidationUsage(wgpu::BufferUsage usage) { - // https://learn.microsoft.com/en-us/windows/win32/api/d3d11/ne-d3d11-d3d11_bind_flag - // D3D11 doesn't support constants buffers with other accelerated GPU usages. - // TODO(dawn:1755): find a way to workaround this D3D11 limitation. - constexpr wgpu::BufferUsage kAllowedUniformBufferUsages = - wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::Uniform; - - DAWN_INVALID_IF( - usage & wgpu::BufferUsage::Uniform && !IsSubset(usage, kAllowedUniformBufferUsages), - "Buffer usage can't be both uniform and other accelerated usages with D3D11"); - - return {}; -} +constexpr wgpu::BufferUsage kD3D11AllowedUniformBufferUsages = + wgpu::BufferUsage::Uniform | wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::CopySrc; // Resource usage Default Dynamic Immutable Staging // ------------------------------------------------------------ @@ -74,19 +63,19 @@ UINT D3D11BufferBindFlags(wgpu::BufferUsage usage) { UINT bindFlags = 0; if (usage & (wgpu::BufferUsage::Vertex)) { - bindFlags |= D3D11_BIND_FLAG::D3D11_BIND_VERTEX_BUFFER; + bindFlags |= D3D11_BIND_VERTEX_BUFFER; } if (usage & wgpu::BufferUsage::Index) { - bindFlags |= D3D11_BIND_FLAG::D3D11_BIND_INDEX_BUFFER; + bindFlags |= D3D11_BIND_INDEX_BUFFER; } if (usage & (wgpu::BufferUsage::Uniform)) { - bindFlags |= D3D11_BIND_FLAG::D3D11_BIND_CONSTANT_BUFFER; + bindFlags |= D3D11_BIND_CONSTANT_BUFFER; } if (usage & (wgpu::BufferUsage::Storage | kInternalStorageBuffer)) { - bindFlags |= D3D11_BIND_FLAG::D3D11_BIND_UNORDERED_ACCESS; + bindFlags |= D3D11_BIND_UNORDERED_ACCESS; } if (usage & kReadOnlyStorageBuffer) { - bindFlags |= D3D11_BIND_FLAG::D3D11_BIND_SHADER_RESOURCE; + bindFlags |= D3D11_BIND_SHADER_RESOURCE; } constexpr wgpu::BufferUsage kCopyUsages = @@ -96,7 +85,7 @@ UINT D3D11BufferBindFlags(wgpu::BufferUsage usage) { // to copy data between buffer and texture. So the buffer needs to be bound as unordered access // view. if (IsSubset(usage, kCopyUsages)) { - bindFlags |= D3D11_BIND_FLAG::D3D11_BIND_UNORDERED_ACCESS; + bindFlags |= D3D11_BIND_UNORDERED_ACCESS; } return bindFlags; @@ -112,8 +101,7 @@ UINT D3D11CpuAccessFlags(wgpu::BufferUsage usage) { // - For texture to buffer copy, we may need copy texture to a staging (mappable) // texture, and then memcpy the data from the staging texture to the staging buffer. So // D3D11_CPU_ACCESS_WRITE is needed to MapRead usage. - cpuAccessFlags = D3D11_CPU_ACCESS_FLAG::D3D11_CPU_ACCESS_READ | - D3D11_CPU_ACCESS_FLAG::D3D11_CPU_ACCESS_WRITE; + cpuAccessFlags = D3D11_CPU_ACCESS_READ | D3D11_CPU_ACCESS_WRITE; } return cpuAccessFlags; } @@ -155,7 +143,6 @@ ResultOrError> Buffer::Create(Device* device, const BufferDescriptor MaybeError Buffer::Initialize(bool mappedAtCreation) { // TODO(dawn:1705): handle mappedAtCreation for NonzeroClearResourcesOnCreationForTesting - DAWN_TRY(ValidationUsage(GetUsage())); // Allocate at least 4 bytes so clamped accesses are always in bounds. uint64_t size = std::max(GetSize(), uint64_t(4u)); @@ -167,19 +154,46 @@ MaybeError Buffer::Initialize(bool mappedAtCreation) { } mAllocatedSize = Align(size, alignment); - // Create mD3d11Buffer - D3D11_BUFFER_DESC bufferDescriptor; - bufferDescriptor.ByteWidth = mAllocatedSize; - bufferDescriptor.Usage = D3D11BufferUsage(GetUsage()); - bufferDescriptor.BindFlags = D3D11BufferBindFlags(GetUsage()); - bufferDescriptor.CPUAccessFlags = D3D11CpuAccessFlags(GetUsage()); - bufferDescriptor.MiscFlags = D3D11BufferMiscFlags(GetUsage()); - bufferDescriptor.StructureByteStride = 0; + bool needsConstantBuffer = GetUsage() & wgpu::BufferUsage::Uniform; + bool onlyNeedsConstantBuffer = + needsConstantBuffer && IsSubset(GetUsage(), kD3D11AllowedUniformBufferUsages); - DAWN_TRY(CheckOutOfMemoryHRESULT(ToBackend(GetDevice()) - ->GetD3D11Device() - ->CreateBuffer(&bufferDescriptor, nullptr, &mD3d11Buffer), - "ID3D11Device::CreateBuffer")); + if (!onlyNeedsConstantBuffer) { + // Create mD3d11NonConstantBuffer + wgpu::BufferUsage nonUniformUsage = GetUsage() & ~wgpu::BufferUsage::Uniform; + D3D11_BUFFER_DESC bufferDescriptor; + bufferDescriptor.ByteWidth = mAllocatedSize; + bufferDescriptor.Usage = D3D11BufferUsage(nonUniformUsage); + bufferDescriptor.BindFlags = D3D11BufferBindFlags(nonUniformUsage); + bufferDescriptor.CPUAccessFlags = D3D11CpuAccessFlags(nonUniformUsage); + bufferDescriptor.MiscFlags = D3D11BufferMiscFlags(nonUniformUsage); + bufferDescriptor.StructureByteStride = 0; + + DAWN_TRY(CheckOutOfMemoryHRESULT( + ToBackend(GetDevice()) + ->GetD3D11Device() + ->CreateBuffer(&bufferDescriptor, nullptr, &mD3d11NonConstantBuffer), + "ID3D11Device::CreateBuffer")); + } + + if (needsConstantBuffer) { + // Create mD3d11ConstantBuffer + D3D11_BUFFER_DESC bufferDescriptor; + bufferDescriptor.ByteWidth = mAllocatedSize; + bufferDescriptor.Usage = D3D11_USAGE_DEFAULT; + bufferDescriptor.BindFlags = D3D11_BIND_CONSTANT_BUFFER; + bufferDescriptor.CPUAccessFlags = 0; + bufferDescriptor.MiscFlags = 0; + bufferDescriptor.StructureByteStride = 0; + + DAWN_TRY(CheckOutOfMemoryHRESULT( + ToBackend(GetDevice()) + ->GetD3D11Device() + ->CreateBuffer(&bufferDescriptor, nullptr, &mD3d11ConstantBuffer), + "ID3D11Device::CreateBuffer")); + } + + ASSERT(mD3d11NonConstantBuffer || mD3d11ConstantBuffer); SetLabelImpl(); return {}; @@ -201,10 +215,11 @@ MaybeError Buffer::MapInternal() { // need write permission to initialize the buffer. // TODO(dawn:1705): investigate the performance impact of mapping with D3D11_MAP_READ_WRITE. D3D11_MAPPED_SUBRESOURCE mappedResource; - DAWN_TRY(CheckHRESULT(commandContext->GetD3D11DeviceContext()->Map( - mD3d11Buffer.Get(), /*Subresource=*/0, D3D11_MAP_READ_WRITE, - /*MapFlags=*/0, &mappedResource), - "ID3D11DeviceContext::Map")); + DAWN_TRY(CheckHRESULT( + commandContext->GetD3D11DeviceContext()->Map(mD3d11NonConstantBuffer.Get(), + /*Subresource=*/0, D3D11_MAP_READ_WRITE, + /*MapFlags=*/0, &mappedResource), + "ID3D11DeviceContext::Map")); mMappedData = reinterpret_cast(mappedResource.pData); return {}; @@ -214,7 +229,8 @@ void Buffer::UnmapInternal() { DAWN_ASSERT(mMappedData); CommandRecordingContext* commandContext = ToBackend(GetDevice())->GetPendingCommandContext(); - commandContext->GetD3D11DeviceContext()->Unmap(mD3d11Buffer.Get(), /*Subresource=*/0); + commandContext->GetD3D11DeviceContext()->Unmap(mD3d11NonConstantBuffer.Get(), + /*Subresource=*/0); mMappedData = nullptr; } @@ -224,7 +240,7 @@ MaybeError Buffer::MapAtCreationImpl() { } MaybeError Buffer::MapAsyncImpl(wgpu::MapMode mode, size_t offset, size_t size) { - DAWN_ASSERT(mD3d11Buffer); + DAWN_ASSERT(mD3d11NonConstantBuffer); // TODO(dawn:1705): make sure the map call is not blocked by the GPU operations. DAWN_TRY(MapInternal()); @@ -236,7 +252,7 @@ MaybeError Buffer::MapAsyncImpl(wgpu::MapMode mode, size_t offset, size_t size) } void Buffer::UnmapImpl() { - DAWN_ASSERT(mD3d11Buffer); + DAWN_ASSERT(mD3d11NonConstantBuffer); DAWN_ASSERT(mMappedData); UnmapInternal(); } @@ -252,11 +268,13 @@ void Buffer::DestroyImpl() { if (mMappedData) { UnmapInternal(); } - mD3d11Buffer = nullptr; + mD3d11NonConstantBuffer = nullptr; } void Buffer::SetLabelImpl() { - SetDebugName(ToBackend(GetDevice()), mD3d11Buffer.Get(), "Dawn_Buffer", GetLabel()); + SetDebugName(ToBackend(GetDevice()), mD3d11NonConstantBuffer.Get(), "Dawn_Buffer", GetLabel()); + SetDebugName(ToBackend(GetDevice()), mD3d11ConstantBuffer.Get(), "Dawn_ConstantBuffer", + GetLabel()); } MaybeError Buffer::EnsureDataInitialized(CommandRecordingContext* commandContext) { @@ -309,6 +327,22 @@ MaybeError Buffer::InitializeToZero(CommandRecordingContext* commandContext) { return {}; } +void Buffer::MarkMutated() { + mConstantBufferIsUpdated = false; +} + +void Buffer::EnsureConstantBufferIsUpdated(CommandRecordingContext* commandContext) { + if (mConstantBufferIsUpdated) { + return; + } + + DAWN_ASSERT(mD3d11NonConstantBuffer); + DAWN_ASSERT(mD3d11ConstantBuffer); + commandContext->GetD3D11DeviceContext1()->CopyResource(mD3d11ConstantBuffer.Get(), + mD3d11NonConstantBuffer.Get()); + mConstantBufferIsUpdated = true; +} + ResultOrError> Buffer::CreateD3D11ShaderResourceView( uint64_t offset, uint64_t size) const { @@ -324,10 +358,11 @@ ResultOrError> Buffer::CreateD3D11ShaderResourc desc.BufferEx.NumElements = numElements; desc.BufferEx.Flags = D3D11_BUFFEREX_SRV_FLAG_RAW; ComPtr srv; - DAWN_TRY(CheckHRESULT(ToBackend(GetDevice()) - ->GetD3D11Device() - ->CreateShaderResourceView(mD3d11Buffer.Get(), &desc, &srv), - "ShaderResourceView creation")); + DAWN_TRY( + CheckHRESULT(ToBackend(GetDevice()) + ->GetD3D11Device() + ->CreateShaderResourceView(mD3d11NonConstantBuffer.Get(), &desc, &srv), + "ShaderResourceView creation")); return srv; } @@ -349,11 +384,11 @@ ResultOrError> Buffer::CreateD3D11UnorderedAc desc.Buffer.Flags = D3D11_BUFFER_UAV_FLAG_RAW; ComPtr uav; - DAWN_TRY(CheckHRESULT(ToBackend(GetDevice()) - ->GetD3D11Device5() - ->CreateUnorderedAccessView1(mD3d11Buffer.Get(), &desc, &uav), - "UnorderedAccessView creation")); - + DAWN_TRY( + CheckHRESULT(ToBackend(GetDevice()) + ->GetD3D11Device5() + ->CreateUnorderedAccessView1(mD3d11NonConstantBuffer.Get(), &desc, &uav), + "UnorderedAccessView creation")); return uav; } @@ -388,6 +423,8 @@ MaybeError Buffer::ClearInternal(CommandRecordingContext* commandContext, if (mMappedData) { memset(mMappedData + offset, clearValue, size); + // The WebGPU uniform buffer is not mappable. + ASSERT(!mD3d11ConstantBuffer); return {}; } @@ -428,6 +465,8 @@ MaybeError Buffer::WriteInternal(CommandRecordingContext* commandContext, if (scopedMap.GetMappedData()) { memcpy(scopedMap.GetMappedData() + offset, data, size); + // The WebGPU uniform buffer is not mappable. + ASSERT(!mD3d11ConstantBuffer); return {}; } @@ -436,65 +475,55 @@ MaybeError Buffer::WriteInternal(CommandRecordingContext* commandContext, ID3D11DeviceContext1* d3d11DeviceContext1 = commandContext->GetD3D11DeviceContext1(); - // For updating the full buffer, just pass nullptr as the pDstBox. - if (offset == 0 && size == GetAllocatedSize()) { - d3d11DeviceContext1->UpdateSubresource(GetD3D11Buffer(), /*DstSubresource=*/0, - /*pDstBox=*/nullptr, data, + if (mD3d11NonConstantBuffer) { + D3D11_BOX box; + box.left = offset; + box.right = offset + size; + box.top = 0; + box.bottom = 1; + box.front = 0; + box.back = 1; + d3d11DeviceContext1->UpdateSubresource(mD3d11NonConstantBuffer.Get(), /*DstSubresource=*/0, + &box, data, /*SrcRowPitch=*/0, /*SrcDepthPitch*/ 0); + if (!mD3d11ConstantBuffer) { + return {}; + } + + // if mConstantBufferIsUpdated is false, the content of mD3d11ConstantBuffer will be + // updated by EnsureConstantBufferIsUpdated() when the constant buffer is about to be used. + if (!mConstantBufferIsUpdated) { + return {}; + } + + // Copy the modified part of the mD3d11NonConstantBuffer to mD3d11ConstantBuffer. + d3d11DeviceContext1->CopySubresourceRegion( + mD3d11ConstantBuffer.Get(), /*DstSubresource=*/0, /*DstX=*/offset, + /*DstY=*/0, + /*DstZ=*/0, mD3d11NonConstantBuffer.Get(), /*SrcSubresource=*/0, &box); + return {}; } - D3D11_BOX box; - box.left = offset; - box.right = offset + size; - box.top = 0; - box.bottom = 1; - box.front = 0; - box.back = 1; + ASSERT(mD3d11ConstantBuffer); - if ((GetUsage() & wgpu::BufferUsage::Uniform)) { - if (!IsAligned(box.left, 16) || !IsAligned(box.right, 16)) { - // Create a temp staging buffer to workaround the alignment issue. - BufferDescriptor descriptor; - descriptor.size = box.right - box.left; - DAWN_ASSERT(IsAligned(descriptor.size, 4)); - descriptor.usage = wgpu::BufferUsage::MapWrite | wgpu::BufferUsage::CopySrc; - descriptor.mappedAtCreation = false; - descriptor.label = "temp staging buffer"; - Ref stagingBufferBase; - DAWN_TRY_ASSIGN(stagingBufferBase, GetDevice()->CreateBuffer(&descriptor)); - Ref stagingBuffer; - stagingBuffer = ToBackend(std::move(stagingBufferBase)); - { - ScopedMap scopedMap; - DAWN_TRY_ASSIGN(scopedMap, ScopedMap::Create(stagingBuffer.Get())); - uint8_t* mappedData = scopedMap.GetMappedData(); - DAWN_ASSERT(mappedData); - memcpy(mappedData, data, size); - } - box.left = 0; - box.right = descriptor.size; - commandContext->GetD3D11DeviceContext()->CopySubresourceRegion( - GetD3D11Buffer(), /*DstSubresource=*/0, /*DstX=*/offset, - /*DstY=*/0, - /*DstZ=*/0, stagingBuffer->GetD3D11Buffer(), /*SrcSubresource=*/0, &box); - stagingBuffer = nullptr; + // If the mD3d11NonConstantBuffer is null, we have to create a staging buffer for transfer the + // data to mD3d11ConstantBuffer, since UpdateSubresource() has many restrictions. For example, + // the size of the data has to be a multiple of 16, etc + BufferDescriptor descriptor; + descriptor.size = size; + DAWN_ASSERT(IsAligned(descriptor.size, 4)); + descriptor.usage = wgpu::BufferUsage::MapWrite | wgpu::BufferUsage::CopySrc; + descriptor.mappedAtCreation = false; + descriptor.label = "DawnWriteStagingBuffer"; + Ref stagingBuffer; + DAWN_TRY_ASSIGN(stagingBuffer, GetDevice()->CreateBuffer(&descriptor)); - } else { - // TODO(dawn:1739): check whether driver supports partial update of uniform buffer. - d3d11DeviceContext1->UpdateSubresource1(GetD3D11Buffer(), /*DstSubresource=*/0, &box, - data, - /*SrcRowPitch=*/0, - /*SrcDepthPitch*/ 0, D3D11_COPY_NO_OVERWRITE); - } - } else { - d3d11DeviceContext1->UpdateSubresource(GetD3D11Buffer(), /*DstSubresource=*/0, &box, data, - /*SrcRowPitch=*/0, - /*SrcDepthPitch*/ 0); - } + DAWN_TRY(ToBackend(stagingBuffer)->WriteInternal(commandContext, 0, data, size)); - return {}; + return Buffer::CopyInternal(commandContext, ToBackend(stagingBuffer.Get()), /*sourceOffset=*/0, + /*size=*/size, this, offset); } // static @@ -509,7 +538,16 @@ MaybeError Buffer::Copy(CommandRecordingContext* commandContext, DAWN_TRY(source->EnsureDataInitialized(commandContext)); DAWN_TRY( destination->EnsureDataInitializedAsDestination(commandContext, destinationOffset, size)); + return CopyInternal(commandContext, source, sourceOffset, size, destination, destinationOffset); +} +// static +MaybeError Buffer::CopyInternal(CommandRecordingContext* commandContext, + Buffer* source, + uint64_t sourceOffset, + size_t size, + Buffer* destination, + uint64_t destinationOffset) { D3D11_BOX srcBox; srcBox.left = sourceOffset; srcBox.right = sourceOffset + size; @@ -517,10 +555,32 @@ MaybeError Buffer::Copy(CommandRecordingContext* commandContext, srcBox.bottom = 1; srcBox.front = 0; srcBox.back = 1; - commandContext->GetD3D11DeviceContext()->CopySubresourceRegion( - destination->mD3d11Buffer.Get(), /*DstSubresource=*/0, /*DstX=*/destinationOffset, - /*DstY=*/0, - /*DstZ=*/0, source->mD3d11Buffer.Get(), /*SrcSubresource=*/0, &srcBox); + ID3D11Buffer* d3d11SourceBuffer = source->mD3d11NonConstantBuffer + ? source->mD3d11NonConstantBuffer.Get() + : source->mD3d11ConstantBuffer.Get(); + ASSERT(d3d11SourceBuffer); + + if (destination->mD3d11NonConstantBuffer) { + commandContext->GetD3D11DeviceContext()->CopySubresourceRegion( + destination->mD3d11NonConstantBuffer.Get(), /*DstSubresource=*/0, + /*DstX=*/destinationOffset, + /*DstY=*/0, + /*DstZ=*/0, d3d11SourceBuffer, /*SrcSubresource=*/0, &srcBox); + } + + // if mConstantBufferIsUpdated is false, the content of mD3d11ConstantBuffer will be + // updated by EnsureConstantBufferIsUpdated() when the constant buffer is about to be used. + if (!destination->mConstantBufferIsUpdated) { + return {}; + } + + if (destination->mD3d11ConstantBuffer) { + commandContext->GetD3D11DeviceContext()->CopySubresourceRegion( + destination->mD3d11ConstantBuffer.Get(), /*DstSubresource=*/0, + /*DstX=*/destinationOffset, + /*DstY=*/0, + /*DstZ=*/0, d3d11SourceBuffer, /*SrcSubresource=*/0, &srcBox); + } return {}; } diff --git a/src/dawn/native/d3d11/BufferD3D11.h b/src/dawn/native/d3d11/BufferD3D11.h index d8633179cb..aac288d817 100644 --- a/src/dawn/native/d3d11/BufferD3D11.h +++ b/src/dawn/native/d3d11/BufferD3D11.h @@ -40,14 +40,19 @@ class Buffer final : public BufferBase { // Dawn API void SetLabelImpl() override; - ID3D11Buffer* GetD3D11Buffer() const { return mD3d11Buffer.Get(); } + ID3D11Buffer* GetD3D11ConstantBuffer() const { return mD3d11ConstantBuffer.Get(); } + ID3D11Buffer* GetD3D11NonConstantBuffer() const { return mD3d11NonConstantBuffer.Get(); } + // Mark the mD3d11NonConstantBuffer is mutated by shaders, if mD3d11ConstantBuffer exists, + // it will be synced with mD3d11NonConstantBuffer before binding it to the constant buffer slot. + void MarkMutated(); + // Update content of the mD3d11ConstantBuffer from mD3d11NonConstantBuffer if needed. + void EnsureConstantBufferIsUpdated(CommandRecordingContext* commandContext); ResultOrError> CreateD3D11ShaderResourceView( uint64_t offset, uint64_t size) const; ResultOrError> CreateD3D11UnorderedAccessView1( uint64_t offset, uint64_t size) const; - MaybeError Clear(CommandRecordingContext* commandContext, uint8_t clearValue, uint64_t offset, @@ -116,9 +121,18 @@ class Buffer final : public BufferBase { uint64_t bufferOffset, const void* data, size_t size); - - // The buffer object can be used as vertex, index, uniform, storage, or indirect buffer. - ComPtr mD3d11Buffer; + // Copy the buffer without checking if the buffer is initialized. + static MaybeError CopyInternal(CommandRecordingContext* commandContext, + Buffer* source, + uint64_t sourceOffset, + size_t size, + Buffer* destination, + uint64_t destinationOffset); + // The buffer object for constant buffer usage. + ComPtr mD3d11ConstantBuffer; + // The buffer object for non-constant buffer usages(e.g. storage buffer, vertex buffer, etc.) + ComPtr mD3d11NonConstantBuffer; + bool mConstantBufferIsUpdated = true; uint8_t* mMappedData = nullptr; }; diff --git a/src/dawn/native/d3d11/CommandBufferD3D11.cpp b/src/dawn/native/d3d11/CommandBufferD3D11.cpp index 1fc3296b55..919e317522 100644 --- a/src/dawn/native/d3d11/CommandBufferD3D11.cpp +++ b/src/dawn/native/d3d11/CommandBufferD3D11.cpp @@ -358,7 +358,7 @@ MaybeError CommandBuffer::ExecuteComputePass(CommandRecordingContext* commandCon } commandContext->GetD3D11DeviceContext()->DispatchIndirect( - indirectBuffer->GetD3D11Buffer(), dispatch->indirectOffset); + indirectBuffer->GetD3D11NonConstantBuffer(), dispatch->indirectOffset); break; } @@ -533,7 +533,7 @@ MaybeError CommandBuffer::ExecuteRenderPass(BeginRenderPassCmd* renderPass, } commandContext->GetD3D11DeviceContext()->DrawInstancedIndirect( - indirectBuffer->GetD3D11Buffer(), draw->indirectOffset); + indirectBuffer->GetD3D11NonConstantBuffer(), draw->indirectOffset); break; } @@ -559,7 +559,7 @@ MaybeError CommandBuffer::ExecuteRenderPass(BeginRenderPassCmd* renderPass, } commandContext->GetD3D11DeviceContext()->DrawIndexedInstancedIndirect( - indirectBuffer->GetD3D11Buffer(), draw->indirectOffset); + indirectBuffer->GetD3D11NonConstantBuffer(), draw->indirectOffset); break; } @@ -594,7 +594,7 @@ MaybeError CommandBuffer::ExecuteRenderPass(BeginRenderPassCmd* renderPass, DXGI_FORMAT indexBufferFormat = DXGIIndexFormat(cmd->format); commandContext->GetD3D11DeviceContext()->IASetIndexBuffer( - ToBackend(cmd->buffer)->GetD3D11Buffer(), indexBufferFormat, + ToBackend(cmd->buffer)->GetD3D11NonConstantBuffer(), indexBufferFormat, indexBufferBaseOffset); break; @@ -602,7 +602,7 @@ MaybeError CommandBuffer::ExecuteRenderPass(BeginRenderPassCmd* renderPass, case Command::SetVertexBuffer: { SetVertexBufferCmd* cmd = iter->NextCommand(); - ID3D11Buffer* buffer = ToBackend(cmd->buffer)->GetD3D11Buffer(); + ID3D11Buffer* buffer = ToBackend(cmd->buffer)->GetD3D11NonConstantBuffer(); vertexBufferTracker.OnSetVertexBuffer(cmd->slot, buffer, cmd->offset); break; } diff --git a/src/dawn/native/d3d11/CommandRecordingContextD3D11.cpp b/src/dawn/native/d3d11/CommandRecordingContextD3D11.cpp index bfbc052731..a7a3c76d58 100644 --- a/src/dawn/native/d3d11/CommandRecordingContextD3D11.cpp +++ b/src/dawn/native/d3d11/CommandRecordingContextD3D11.cpp @@ -63,7 +63,7 @@ MaybeError CommandRecordingContext::Open(Device* device) { // Always bind the uniform buffer to the reserved slot for all pipelines. // This buffer will be updated with the correct values before each draw or dispatch call. - ID3D11Buffer* bufferPtr = mUniformBuffer->GetD3D11Buffer(); + ID3D11Buffer* bufferPtr = mUniformBuffer->GetD3D11ConstantBuffer(); mD3D11DeviceContext4->VSSetConstantBuffers(PipelineLayout::kReservedConstantBufferSlot, 1, &bufferPtr); mD3D11DeviceContext4->CSSetConstantBuffers(PipelineLayout::kReservedConstantBufferSlot, 1, diff --git a/src/dawn/native/d3d11/TextureD3D11.h b/src/dawn/native/d3d11/TextureD3D11.h index 06e84a21ca..ecf08c44d1 100644 --- a/src/dawn/native/d3d11/TextureD3D11.h +++ b/src/dawn/native/d3d11/TextureD3D11.h @@ -61,7 +61,6 @@ class Texture final : public d3d::Texture { D3D11_DEPTH_STENCIL_VIEW_DESC GetDSVDescriptor(const SubresourceRange& range, bool depthReadOnly, bool stencilReadOnly) const; - MaybeError EnsureSubresourceContentInitialized(CommandRecordingContext* commandContext, const SubresourceRange& range); diff --git a/src/dawn/tests/BUILD.gn b/src/dawn/tests/BUILD.gn index c2735ba224..0d387472b1 100644 --- a/src/dawn/tests/BUILD.gn +++ b/src/dawn/tests/BUILD.gn @@ -664,6 +664,10 @@ source_set("white_box_tests_sources") { "white_box/QueryInternalShaderTests.cpp", ] + if (dawn_enable_d3d11) { + sources += [ "white_box/D3D11BufferTests.cpp" ] + } + if (dawn_enable_d3d12) { sources += [ "white_box/D3D12DescriptorHeapTests.cpp", diff --git a/src/dawn/tests/white_box/D3D11BufferTests.cpp b/src/dawn/tests/white_box/D3D11BufferTests.cpp new file mode 100644 index 0000000000..4ece2ef9d4 --- /dev/null +++ b/src/dawn/tests/white_box/D3D11BufferTests.cpp @@ -0,0 +1,312 @@ +// Copyright 2023 The Dawn Authors +// +// Licensed under the Apache License, Version 2.0 (the "License"); +// you may not use this file except in compliance with the License. +// You may obtain a copy of the License at +// +// http://www.apache.org/licenses/LICENSE-2.0 +// +// Unless required by applicable law or agreed to in writing, software +// distributed under the License is distributed on an "AS IS" BASIS, +// WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. +// See the License for the specific language governing permissions and +// limitations under the License. + +#include + +#include "dawn/native/D3D11Backend.h" +#include "dawn/native/d3d11/BufferD3D11.h" +#include "dawn/native/d3d11/DeviceD3D11.h" +#include "dawn/tests/DawnTest.h" +#include "dawn/utils/ComboRenderPipelineDescriptor.h" +#include "dawn/utils/WGPUHelpers.h" + +class D3D11BufferTests : public DawnTest { + protected: + void SetUp() override { + DawnTest::SetUp(); + DAWN_TEST_UNSUPPORTED_IF(UsesWire()); + } + + wgpu::Buffer CreateBuffer(uint32_t bufferSize, wgpu::BufferUsage usage) { + wgpu::BufferDescriptor descriptor; + + descriptor.size = bufferSize; + descriptor.usage = usage; + + return device.CreateBuffer(&descriptor); + } + + ID3D11Device* GetD3D11Device() { + return dawn::native::d3d11::ToBackend(dawn::native::FromAPI((device.Get()))) + ->GetD3D11Device(); + } + + template + void CheckBuffer(ID3D11Buffer* buffer, std::vector expectedData, uint32_t offset = 0) { + D3D11_BUFFER_DESC bufferDesc; + buffer->GetDesc(&bufferDesc); + EXPECT_GE(bufferDesc.ByteWidth, (expectedData.size() + offset) * sizeof(T)); + + // Create D3D11 staging buffer + D3D11_BUFFER_DESC desc; + desc.ByteWidth = expectedData.size() * sizeof(T); + desc.Usage = D3D11_USAGE_STAGING; + desc.BindFlags = 0; + desc.CPUAccessFlags = D3D11_CPU_ACCESS_READ; + desc.MiscFlags = 0; + desc.StructureByteStride = 0; + + ComPtr stagingBuffer; + ASSERT_HRESULT_SUCCEEDED(GetD3D11Device()->CreateBuffer(&desc, nullptr, &stagingBuffer)); + + ID3D11DeviceContext* deviceContext; + GetD3D11Device()->GetImmediateContext(&deviceContext); + + // Copy buffer to staging buffer + D3D11_BOX srcBox; + srcBox.left = offset * sizeof(T); + srcBox.right = (offset + expectedData.size()) * sizeof(T); + srcBox.top = 0; + srcBox.bottom = 1; + srcBox.front = 0; + srcBox.back = 1; + deviceContext->CopySubresourceRegion(stagingBuffer.Get(), 0, 0, 0, 0, buffer, 0, &srcBox); + + // Map staging buffer + D3D11_MAPPED_SUBRESOURCE mappedResource; + ASSERT_HRESULT_SUCCEEDED( + deviceContext->Map(stagingBuffer.Get(), 0, D3D11_MAP_READ, 0, &mappedResource)); + + // Check data + const T* actualData = reinterpret_cast(mappedResource.pData); + for (size_t i = 0; i < expectedData.size(); ++i) { + EXPECT_EQ(expectedData[i], actualData[i]); + } + + // Unmap staging buffer + deviceContext->Unmap(stagingBuffer.Get(), 0); + } +}; + +// Test that creating a uniform buffer +TEST_P(D3D11BufferTests, CreateUniformBuffer) { + { + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_EQ(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } + { + wgpu::BufferUsage usage = + wgpu::BufferUsage::Uniform | wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::CopySrc; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_EQ(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } + { + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::Vertex; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } + { + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::Index; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } + { + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::Indirect; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } + { + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::Storage; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } + { + wgpu::BufferUsage usage = wgpu::BufferUsage::Storage; + wgpu::Buffer buffer = CreateBuffer(4, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_EQ(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + } +} + +// Test Buffer::Write() +TEST_P(D3D11BufferTests, WriteUniformBuffer) { + { + std::vector data = {0x12, 0x34, 0x56, 0x78}; + wgpu::BufferUsage usage = + wgpu::BufferUsage::Uniform | wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::CopySrc; + wgpu::Buffer buffer = CreateBuffer(data.size(), usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_EQ(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + + queue.WriteBuffer(buffer, 0, data.data(), data.size()); + EXPECT_BUFFER_U8_RANGE_EQ(data.data(), buffer, 0, data.size()); + + CheckBuffer(d3d11Buffer->GetD3D11ConstantBuffer(), data); + } + { + std::vector data = {0x12, 0x34, 0x56, 0x78}; + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::Vertex | + wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::CopySrc; + wgpu::Buffer buffer = CreateBuffer(data.size(), usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + + queue.WriteBuffer(buffer, 0, data.data(), data.size()); + EXPECT_BUFFER_U8_RANGE_EQ(data.data(), buffer, 0, data.size()); + + // both buffers should be updated. + CheckBuffer(d3d11Buffer->GetD3D11NonConstantBuffer(), data); + CheckBuffer(d3d11Buffer->GetD3D11ConstantBuffer(), data); + } +} + +// Test UAV write +TEST_P(D3D11BufferTests, WriteUniformBufferWithComputeShader) { + constexpr size_t kNumValues = 100; + std::vector data(kNumValues, 0x12345678); + uint64_t bufferSize = static_cast(data.size() * sizeof(uint32_t)); + wgpu::BufferUsage usage = wgpu::BufferUsage::Uniform | wgpu::BufferUsage::Storage | + wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::CopySrc; + wgpu::Buffer buffer = CreateBuffer(bufferSize, usage); + dawn::native::d3d11::Buffer* d3d11Buffer = + dawn::native::d3d11::ToBackend(dawn::native::FromAPI(buffer.Get())); + + EXPECT_NE(d3d11Buffer->GetD3D11NonConstantBuffer(), nullptr); + EXPECT_NE(d3d11Buffer->GetD3D11ConstantBuffer(), nullptr); + + queue.WriteBuffer(buffer, 0, data.data(), bufferSize); + EXPECT_BUFFER_U32_RANGE_EQ(data.data(), buffer, 0, data.size()); + + CheckBuffer(d3d11Buffer->GetD3D11NonConstantBuffer(), data); + CheckBuffer(d3d11Buffer->GetD3D11ConstantBuffer(), data); + + // Fill the buffer with 0x11223344 with a compute shader + { + wgpu::ShaderModule module = utils::CreateShaderModule(device, R"( + struct Buf { + data : array + } + + @group(0) @binding(0) var buf : Buf; + + @compute @workgroup_size(1) + fn main(@builtin(global_invocation_id) GlobalInvocationID : vec3u) { + buf.data[GlobalInvocationID.x] = + vec4u(0x11223344u, 0x11223344u, 0x11223344u, 0x11223344u); + } + )"); + + wgpu::ComputePipelineDescriptor pipelineDesc = {}; + pipelineDesc.compute.module = module; + pipelineDesc.compute.entryPoint = "main"; + wgpu::ComputePipeline pipeline = device.CreateComputePipeline(&pipelineDesc); + + wgpu::BindGroup bindGroupA = utils::MakeBindGroup(device, pipeline.GetBindGroupLayout(0), + { + {0, buffer, 0, bufferSize}, + }); + + wgpu::CommandEncoder encoder = device.CreateCommandEncoder(); + wgpu::ComputePassEncoder pass = encoder.BeginComputePass(); + pass.SetPipeline(pipeline); + pass.SetBindGroup(0, bindGroupA); + pass.DispatchWorkgroups(kNumValues / 4); + pass.End(); + + wgpu::CommandBuffer commands = encoder.Finish(); + queue.Submit(1, &commands); + + std::vector expectedData(kNumValues, 0x11223344); + EXPECT_BUFFER_U32_RANGE_EQ(expectedData.data(), buffer, 0, expectedData.size()); + // The non-constant buffer should be updated. + CheckBuffer(d3d11Buffer->GetD3D11NonConstantBuffer(), expectedData); + // The constant buffer should not be updated, until the constant buffer is used a pipeline + CheckBuffer(d3d11Buffer->GetD3D11ConstantBuffer(), data); + } + + // Copy the uniform buffer content to a new buffer with Compute shader + { + wgpu::Buffer newBuffer = + CreateBuffer(bufferSize, wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc); + wgpu::ShaderModule module = utils::CreateShaderModule(device, R"( + struct Buf { + data : array + } + + @group(0) @binding(0) var src : Buf; + @group(0) @binding(1) var dst : Buf; + + @compute @workgroup_size(1) + fn main(@builtin(global_invocation_id) GlobalInvocationID : vec3u) { + dst.data[GlobalInvocationID.x] = src.data[GlobalInvocationID.x]; + } + )"); + + wgpu::ComputePipelineDescriptor pipelineDesc = {}; + pipelineDesc.compute.module = module; + pipelineDesc.compute.entryPoint = "main"; + wgpu::ComputePipeline pipeline = device.CreateComputePipeline(&pipelineDesc); + + wgpu::BindGroup bindGroupA = utils::MakeBindGroup(device, pipeline.GetBindGroupLayout(0), + { + {0, buffer, 0, bufferSize}, + {1, newBuffer, 0, bufferSize}, + }); + + wgpu::CommandEncoder encoder = device.CreateCommandEncoder(); + wgpu::ComputePassEncoder pass = encoder.BeginComputePass(); + pass.SetPipeline(pipeline); + pass.SetBindGroup(0, bindGroupA); + pass.DispatchWorkgroups(kNumValues / 4); + pass.End(); + + wgpu::CommandBuffer commands = encoder.Finish(); + queue.Submit(1, &commands); + + std::vector expectedData(kNumValues, 0x11223344); + EXPECT_BUFFER_U32_RANGE_EQ(expectedData.data(), buffer, 0, expectedData.size()); + EXPECT_BUFFER_U32_RANGE_EQ(expectedData.data(), newBuffer, 0, expectedData.size()); + + // The non-constant buffer should be updated. + CheckBuffer(d3d11Buffer->GetD3D11NonConstantBuffer(), expectedData); + // The constant buffer should be updated too. + CheckBuffer(d3d11Buffer->GetD3D11ConstantBuffer(), expectedData); + } +} + +DAWN_INSTANTIATE_TEST(D3D11BufferTests, D3D11Backend());