d3d11: support uniform buffer with other GPU accelerated usage
d3d11 doesn't allow creating a constant buffer with other GPU
accelerated usage. This CL workarounds problem by creating two
buffers one for uniform buffer usage, one for other usage, and
copy content to uniform buffer when it is needed.
Bug: dawn:1755
Bug: dawn:1798
Bug: dawn:1721
Change-Id: I26bfee1cca2204f6464ba611872c490165e97f68
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/132020
Commit-Queue: Peng Huang <penghuang@chromium.org>
Reviewed-by: Austin Eng <enga@chromium.org>
Kokoro: Kokoro <noreply+kokoro@google.com>
diff --git a/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp b/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp
index b7628cd..8fd2d60 100644
--- a/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp
+++ b/src/dawn/native/d3d11/BindGroupTrackerD3D11.cpp
@@ -132,7 +132,6 @@
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 @@
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 @@
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 b399870..233c195 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 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 @@
// 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 @@
// - 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 @@
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 @@
}
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 @@
// 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<uint8_t*>(mappedResource.pData);
return {};
@@ -214,7 +229,8 @@
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::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 @@
}
void Buffer::UnmapImpl() {
- DAWN_ASSERT(mD3d11Buffer);
+ DAWN_ASSERT(mD3d11NonConstantBuffer);
DAWN_ASSERT(mMappedData);
UnmapInternal();
}
@@ -252,11 +268,13 @@
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 @@
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<ComPtr<ID3D11ShaderResourceView>> Buffer::CreateD3D11ShaderResourceView(
uint64_t offset,
uint64_t size) const {
@@ -324,10 +358,11 @@
desc.BufferEx.NumElements = numElements;
desc.BufferEx.Flags = D3D11_BUFFEREX_SRV_FLAG_RAW;
ComPtr<ID3D11ShaderResourceView> 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 @@
desc.Buffer.Flags = D3D11_BUFFER_UAV_FLAG_RAW;
ComPtr<ID3D11UnorderedAccessView1> 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 @@
if (mMappedData) {
memset(mMappedData + offset, clearValue, size);
+ // The WebGPU uniform buffer is not mappable.
+ ASSERT(!mD3d11ConstantBuffer);
return {};
}
@@ -428,6 +465,8 @@
if (scopedMap.GetMappedData()) {
memcpy(scopedMap.GetMappedData() + offset, data, size);
+ // The WebGPU uniform buffer is not mappable.
+ ASSERT(!mD3d11ConstantBuffer);
return {};
}
@@ -436,65 +475,55 @@
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<BufferBase> stagingBufferBase;
- DAWN_TRY_ASSIGN(stagingBufferBase, GetDevice()->CreateBuffer(&descriptor));
- Ref<Buffer> 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<BufferBase> 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 @@
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 @@
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 d863317..aac288d 100644
--- a/src/dawn/native/d3d11/BufferD3D11.h
+++ b/src/dawn/native/d3d11/BufferD3D11.h
@@ -40,14 +40,19 @@
// 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<ComPtr<ID3D11ShaderResourceView>> CreateD3D11ShaderResourceView(
uint64_t offset,
uint64_t size) const;
ResultOrError<ComPtr<ID3D11UnorderedAccessView1>> CreateD3D11UnorderedAccessView1(
uint64_t offset,
uint64_t size) const;
-
MaybeError Clear(CommandRecordingContext* commandContext,
uint8_t clearValue,
uint64_t offset,
@@ -116,9 +121,18 @@
uint64_t bufferOffset,
const void* data,
size_t size);
-
- // The buffer object can be used as vertex, index, uniform, storage, or indirect buffer.
- ComPtr<ID3D11Buffer> 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<ID3D11Buffer> mD3d11ConstantBuffer;
+ // The buffer object for non-constant buffer usages(e.g. storage buffer, vertex buffer, etc.)
+ ComPtr<ID3D11Buffer> 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 1fc3296..919e317 100644
--- a/src/dawn/native/d3d11/CommandBufferD3D11.cpp
+++ b/src/dawn/native/d3d11/CommandBufferD3D11.cpp
@@ -358,7 +358,7 @@
}
commandContext->GetD3D11DeviceContext()->DispatchIndirect(
- indirectBuffer->GetD3D11Buffer(), dispatch->indirectOffset);
+ indirectBuffer->GetD3D11NonConstantBuffer(), dispatch->indirectOffset);
break;
}
@@ -533,7 +533,7 @@
}
commandContext->GetD3D11DeviceContext()->DrawInstancedIndirect(
- indirectBuffer->GetD3D11Buffer(), draw->indirectOffset);
+ indirectBuffer->GetD3D11NonConstantBuffer(), draw->indirectOffset);
break;
}
@@ -559,7 +559,7 @@
}
commandContext->GetD3D11DeviceContext()->DrawIndexedInstancedIndirect(
- indirectBuffer->GetD3D11Buffer(), draw->indirectOffset);
+ indirectBuffer->GetD3D11NonConstantBuffer(), draw->indirectOffset);
break;
}
@@ -594,7 +594,7 @@
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 @@
case Command::SetVertexBuffer: {
SetVertexBufferCmd* cmd = iter->NextCommand<SetVertexBufferCmd>();
- 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 bfbc052..a7a3c76 100644
--- a/src/dawn/native/d3d11/CommandRecordingContextD3D11.cpp
+++ b/src/dawn/native/d3d11/CommandRecordingContextD3D11.cpp
@@ -63,7 +63,7 @@
// 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 06e84a2..ecf08c4 100644
--- a/src/dawn/native/d3d11/TextureD3D11.h
+++ b/src/dawn/native/d3d11/TextureD3D11.h
@@ -61,7 +61,6 @@
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 c2735ba..0d38747 100644
--- a/src/dawn/tests/BUILD.gn
+++ b/src/dawn/tests/BUILD.gn
@@ -664,6 +664,10 @@
"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 0000000..4ece2ef9
--- /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 <vector>
+
+#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 <typename T>
+ void CheckBuffer(ID3D11Buffer* buffer, std::vector<T> 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<ID3D11Buffer> 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<const T*>(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<uint8_t> 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<uint8_t> 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<uint32_t> data(kNumValues, 0x12345678);
+ uint64_t bufferSize = static_cast<uint64_t>(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<vec4u, 25>
+ }
+
+ @group(0) @binding(0) var<storage, read_write> 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<uint32_t> 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<vec4u, 25>
+ }
+
+ @group(0) @binding(0) var<uniform> src : Buf;
+ @group(0) @binding(1) var<storage, read_write> 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<uint32_t> 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());
