Add SharedBufferMemory Begin/EndAccess Implementation Adds Begin/EndAccess implementations for SharedBufferMemory. Includes generic and D3D12 tests. Bug: dawn:2382 Change-Id: I76a72ea8e8e2296ffc2fddbd9f8b17b22c558f63 Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/180341 Reviewed-by: Austin Eng <enga@chromium.org> Commit-Queue: Brandon1 Jones <brandon1.jones@intel.com>
diff --git a/docs/dawn/features/shared_buffer_memory.md b/docs/dawn/features/shared_buffer_memory.md index fa6b55c..5cfb386 100644 --- a/docs/dawn/features/shared_buffer_memory.md +++ b/docs/dawn/features/shared_buffer_memory.md
@@ -2,7 +2,9 @@ ## Overview -TODO(dawn:2382): Add feature name(s) when implementation is complete. +Shared Buffer Memory refers to a superset of features that allow Dawn to import externally allocated buffers. + +- `wgpu::FeatureName::SharedBufferMemoryD3D12Resource` ```c++ wgpu::SharedBufferMemoryFooBarDescriptor fooBarDesc = { @@ -61,6 +63,10 @@ memory.EndAccess(buffer, &endAccessDesc); ``` +# Uniform Usage Restriction + +Using wgpu:BufferUsage::Uniform with a buffer created from SharedBufferMemory is not allowed due to an alignment restriction on D3D12 when creating a constant buffer view. It is possible this restriction could be removed in the future if additional alignment restrictions are placed on the provided SharedBufferMemory during import. + # Mappable Buffers A buffer created from shared buffer memory cannot be mapped until access to the memory is explicitly started using `BeginAccess`. The buffer must be unmapped before calling `EndAccess`.
diff --git a/src/dawn/dawn.json b/src/dawn/dawn.json index e777ef9..a8e97e3 100644 --- a/src/dawn/dawn.json +++ b/src/dawn/dawn.json
@@ -1867,7 +1867,7 @@ "tags": ["dawn", "native"], "members": [ {"name": "initialized", "type": "bool"}, - {"name": "fence count", "type": "size_t"}, + {"name": "fence count", "type": "size_t", "default": "0"}, {"name": "fences", "type": "shared fence", "annotation": "const*", "length": "fence count"}, {"name": "signaled values", "type": "uint64_t", "annotation": "const*", "length": "fence count"} ] @@ -1878,7 +1878,7 @@ "tags": ["dawn", "native"], "members": [ {"name": "initialized", "type": "bool"}, - {"name": "fence count", "type": "size_t"}, + {"name": "fence count", "type": "size_t", "default": "0"}, {"name": "fences", "type": "shared fence", "annotation": "const*", "length": "fence count"}, {"name": "signaled values", "type": "uint64_t", "annotation": "const*", "length": "fence count"} ]
diff --git a/src/dawn/native/Buffer.cpp b/src/dawn/native/Buffer.cpp index d7702b4..b1084d4 100644 --- a/src/dawn/native/Buffer.cpp +++ b/src/dawn/native/Buffer.cpp
@@ -764,7 +764,7 @@ case BufferState::HostMappedPersistent: return DAWN_VALIDATION_ERROR("Host-mapped %s cannot be mapped again.", this); case BufferState::SharedMemoryNoAccess: - return DAWN_VALIDATION_ERROR("%s used in submit without shared memory access.", this); + return DAWN_VALIDATION_ERROR("%s used without shared memory access.", this); case BufferState::Unmapped: break; }
diff --git a/src/dawn/native/SharedBufferMemory.cpp b/src/dawn/native/SharedBufferMemory.cpp index cc9bded..21a5697 100644 --- a/src/dawn/native/SharedBufferMemory.cpp +++ b/src/dawn/native/SharedBufferMemory.cpp
@@ -32,6 +32,7 @@ #include "dawn/native/Buffer.h" #include "dawn/native/ChainUtils.h" #include "dawn/native/Device.h" +#include "dawn/native/Queue.h" namespace dawn::native { @@ -42,10 +43,20 @@ ErrorSharedBufferMemory(DeviceBase* device, const SharedBufferMemoryDescriptor* descriptor) : SharedBufferMemoryBase(device, descriptor, ObjectBase::kError) {} + Ref<SharedResourceMemoryContents> CreateContents() override { DAWN_UNREACHABLE(); } ResultOrError<Ref<BufferBase>> CreateBufferImpl( const UnpackedPtr<BufferDescriptor>& descriptor) override { DAWN_UNREACHABLE(); } + MaybeError BeginAccessImpl(BufferBase* buffer, + const UnpackedPtr<BeginAccessDescriptor>& descriptor) override { + DAWN_UNREACHABLE(); + } + ResultOrError<FenceAndSignalValue> EndAccessImpl(BufferBase* buffer, + UnpackedPtr<EndAccessState>& state) override { + DAWN_UNREACHABLE(); + } + void DestroyImpl() override {} }; } // namespace @@ -113,6 +124,12 @@ UnpackedPtr<BufferDescriptor> descriptor; DAWN_TRY_ASSIGN(descriptor, ValidateBufferDescriptor(GetDevice(), rawDescriptor)); + // Emit a specific error message if the user attempts to create a buffer with Uniform usage. + DAWN_INVALID_IF(descriptor->usage & wgpu::BufferUsage::Uniform, + "The buffer usage (%s) contains (%s), which is not allowed on buffers created " + "from SharedBufferMemory.", + descriptor->usage, wgpu::BufferUsage::Uniform); + // Ensure the buffer descriptor usage is a subset of the shared buffer memory's usage. DAWN_INVALID_IF(!IsSubset(descriptor->usage, mProperties.usage), "The buffer usage (%s) is incompatible with the SharedBufferMemory usage (%s).",
diff --git a/src/dawn/native/SharedBufferMemory.h b/src/dawn/native/SharedBufferMemory.h index 64c1731..f8e60e1 100644 --- a/src/dawn/native/SharedBufferMemory.h +++ b/src/dawn/native/SharedBufferMemory.h
@@ -65,13 +65,13 @@ const SharedBufferMemoryDescriptor* descriptor, ObjectBase::ErrorTag tag); - SharedBufferMemoryProperties mProperties; - private: ResultOrError<Ref<BufferBase>> CreateBuffer(const BufferDescriptor* rawDescriptor); virtual ResultOrError<Ref<BufferBase>> CreateBufferImpl( const UnpackedPtr<BufferDescriptor>& descriptor) = 0; + + SharedBufferMemoryProperties mProperties; }; } // namespace dawn::native
diff --git a/src/dawn/native/SharedResourceMemory.cpp b/src/dawn/native/SharedResourceMemory.cpp index 81eb653..334c792 100644 --- a/src/dawn/native/SharedResourceMemory.cpp +++ b/src/dawn/native/SharedResourceMemory.cpp
@@ -54,11 +54,11 @@ void SharedResourceMemory::DestroyImpl() {} bool SharedResourceMemory::HasWriteAccess() const { - return mHasWriteAccess; + return mSharedResourceAccessState == SharedResourceAccessState::Write; } bool SharedResourceMemory::HasExclusiveReadAccess() const { - return mHasExclusiveReadAccess; + return mSharedResourceAccessState == SharedResourceAccessState::ExclusiveRead; } int SharedResourceMemory::GetReadAccessCount() const { @@ -125,22 +125,25 @@ DAWN_INVALID_IF(resource->GetFormat().IsMultiPlanar() && !descriptor->initialized, "%s with multiplanar format (%s) must be initialized.", resource, resource->GetFormat().format); - } - DAWN_INVALID_IF(mHasWriteAccess, "%s is currently accessed for writing.", this); - DAWN_INVALID_IF(mHasExclusiveReadAccess, "%s is currently accessed for exclusive reading.", - this); - if constexpr (std::is_same_v<Resource, TextureBase>) { + DAWN_INVALID_IF(HasWriteAccess(), "%s is currently accessed for writing.", this); + DAWN_INVALID_IF(HasExclusiveReadAccess(), "%s is currently accessed for exclusive reading.", + this); + if (static_cast<TextureBase*>(resource)->IsReadOnly()) { if (descriptor->concurrentRead) { + DAWN_ASSERT(!mExclusiveAccess); DAWN_INVALID_IF(!descriptor->initialized, "Concurrent reading an uninitialized %s.", resource); ++mReadAccessCount; + mSharedResourceAccessState = SharedResourceAccessState::SimultaneousRead; + } else { DAWN_INVALID_IF( mReadAccessCount != 0, "Exclusive read access used while %s is currently accessed for reading.", this); - mHasExclusiveReadAccess = true; + mSharedResourceAccessState = SharedResourceAccessState::ExclusiveRead; + mExclusiveAccess = resource; } } else { DAWN_INVALID_IF(descriptor->concurrentRead, "Concurrent reading read-write %s.", @@ -148,8 +151,15 @@ DAWN_INVALID_IF(mReadAccessCount != 0, "Read-Write access used while %s is currently accessed for reading.", this); - mHasWriteAccess = true; + mSharedResourceAccessState = SharedResourceAccessState::Write; + mExclusiveAccess = resource; } + } else if constexpr (std::is_same_v<Resource, BufferBase>) { + DAWN_INVALID_IF(mExclusiveAccess != nullptr, + "Cannot begin access with %s on %s which is currently accessed by %s.", + resource, this, mExclusiveAccess.Get()); + mSharedResourceAccessState = SharedResourceAccessState::Write; + mExclusiveAccess = resource; } DAWN_TRY(BeginAccessImpl(resource, descriptor)); @@ -219,20 +229,35 @@ DAWN_INVALID_IF(!resource->HasAccess(), "%s is not currently being accessed.", resource); if constexpr (std::is_same_v<Resource, TextureBase>) { if (static_cast<TextureBase*>(resource)->IsReadOnly()) { - DAWN_ASSERT(!mHasWriteAccess); - if (mHasExclusiveReadAccess) { + DAWN_ASSERT(!HasWriteAccess()); + if (HasExclusiveReadAccess()) { DAWN_ASSERT(mReadAccessCount == 0); - mHasExclusiveReadAccess = false; + mSharedResourceAccessState = SharedResourceAccessState::NotAccessed; + mExclusiveAccess = nullptr; } else { - DAWN_ASSERT(!mHasExclusiveReadAccess); + DAWN_ASSERT(mSharedResourceAccessState == + SharedResourceAccessState::SimultaneousRead); + DAWN_ASSERT(mExclusiveAccess == nullptr); --mReadAccessCount; + if (mReadAccessCount == 0) { + mSharedResourceAccessState = SharedResourceAccessState::NotAccessed; + } } } else { - DAWN_ASSERT(mHasWriteAccess); - DAWN_ASSERT(!mHasExclusiveReadAccess); + DAWN_ASSERT(mSharedResourceAccessState == SharedResourceAccessState::Write); DAWN_ASSERT(mReadAccessCount == 0); - mHasWriteAccess = false; + mSharedResourceAccessState = SharedResourceAccessState::NotAccessed; + mExclusiveAccess = nullptr; } + } else if constexpr (std::is_same_v<Resource, BufferBase>) { + DAWN_INVALID_IF( + static_cast<BufferBase*>(resource)->APIGetMapState() != wgpu::BufferMapState::Unmapped, + "%s is currently mapped or pending map.", resource); + DAWN_INVALID_IF(mExclusiveAccess != resource, + "Cannot end access with %s on %s which is currently accessed by %s.", + resource, this, mExclusiveAccess.Get()); + mSharedResourceAccessState = SharedResourceAccessState::NotAccessed; + mExclusiveAccess = nullptr; } PendingFenceList fenceList;
diff --git a/src/dawn/native/SharedResourceMemory.h b/src/dawn/native/SharedResourceMemory.h index a587dbe..64ec58e 100644 --- a/src/dawn/native/SharedResourceMemory.h +++ b/src/dawn/native/SharedResourceMemory.h
@@ -42,6 +42,8 @@ class SharedResourceMemoryContents; +enum SharedResourceAccessState { NotAccessed, ExclusiveRead, SimultaneousRead, Write }; + class SharedResource : public ApiObjectBase { public: using ApiObjectBase::ApiObjectBase; @@ -133,8 +135,8 @@ BufferBase* buffer, UnpackedPtr<SharedBufferMemoryEndAccessState>& state); - bool mHasWriteAccess = false; - bool mHasExclusiveReadAccess = false; + Ref<SharedResource> mExclusiveAccess; + SharedResourceAccessState mSharedResourceAccessState = SharedResourceAccessState::NotAccessed; int mReadAccessCount = 0; Ref<SharedResourceMemoryContents> mContents; };
diff --git a/src/dawn/native/d3d12/BufferD3D12.cpp b/src/dawn/native/d3d12/BufferD3D12.cpp index a8075ba..8a6a1a0 100644 --- a/src/dawn/native/d3d12/BufferD3D12.cpp +++ b/src/dawn/native/d3d12/BufferD3D12.cpp
@@ -428,8 +428,10 @@ // evicted. This buffer should already have been made resident when it was created. TRACE_EVENT0(GetDevice()->GetPlatform(), General, "BufferD3D12::MapInternal"); - Heap* heap = ToBackend(mResourceAllocation.GetResourceHeap()); - DAWN_TRY(ToBackend(GetDevice())->GetResidencyManager()->LockAllocation(heap)); + if (mResourceAllocation.GetInfo().mMethod != AllocationMethod::kExternal) { + Heap* heap = ToBackend(mResourceAllocation.GetResourceHeap()); + DAWN_TRY(ToBackend(GetDevice())->GetResidencyManager()->LockAllocation(heap)); + } D3D12_RANGE range = {offset, offset + size}; // mMappedData is the pointer to the start of the resource, irrespective of offset. @@ -481,8 +483,10 @@ // When buffers are mapped, they are locked to keep them in resident memory. We must unlock // them when they are unmapped. - Heap* heap = ToBackend(mResourceAllocation.GetResourceHeap()); - ToBackend(GetDevice())->GetResidencyManager()->UnlockAllocation(heap); + if (mResourceAllocation.GetInfo().mMethod != AllocationMethod::kExternal) { + Heap* heap = ToBackend(mResourceAllocation.GetResourceHeap()); + ToBackend(GetDevice())->GetResidencyManager()->UnlockAllocation(heap); + } } void* Buffer::GetMappedPointer() {
diff --git a/src/dawn/native/d3d12/QueueD3D12.h b/src/dawn/native/d3d12/QueueD3D12.h index 0b2e027..22a88fc 100644 --- a/src/dawn/native/d3d12/QueueD3D12.h +++ b/src/dawn/native/d3d12/QueueD3D12.h
@@ -52,6 +52,7 @@ MaybeError WaitForSerial(ExecutionSerial serial); CommandRecordingContext* GetPendingCommandContext(SubmitMode submitMode = SubmitMode::Normal); ID3D12CommandQueue* GetCommandQueue() const; + ResultOrError<Ref<d3d::SharedFence>> GetOrCreateSharedFence() override; ID3D12SharingContract* GetSharingContract() const; MaybeError SubmitPendingCommands() override; @@ -68,7 +69,6 @@ void ForceEventualFlushOfCommands() override; MaybeError WaitForIdleForDestruction() override; - ResultOrError<Ref<d3d::SharedFence>> GetOrCreateSharedFence() override; void SetEventOnCompletion(ExecutionSerial serial, HANDLE event) override; MaybeError OpenPendingCommands();
diff --git a/src/dawn/native/d3d12/SharedBufferMemoryD3D12.cpp b/src/dawn/native/d3d12/SharedBufferMemoryD3D12.cpp index a242520..005a5d1 100644 --- a/src/dawn/native/d3d12/SharedBufferMemoryD3D12.cpp +++ b/src/dawn/native/d3d12/SharedBufferMemoryD3D12.cpp
@@ -25,11 +25,18 @@ // OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE // OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. +#include "dawn/native/d3d12/SharedBufferMemoryD3D12.h" + #include <utility> +#include "dawn/native/Buffer.h" +#include "dawn/native/ChainUtils.h" +#include "dawn/native/d3d/D3DError.h" +#include "dawn/native/d3d/SharedFenceD3D.h" +#include "dawn/native/d3d/UtilsD3D.h" #include "dawn/native/d3d12/BufferD3D12.h" #include "dawn/native/d3d12/DeviceD3D12.h" -#include "dawn/native/d3d12/SharedBufferMemoryD3D12.h" +#include "dawn/native/d3d12/QueueD3D12.h" namespace dawn::native::d3d12 { @@ -39,6 +46,10 @@ ComPtr<ID3D12Resource> resource) : SharedBufferMemoryBase(device, label, properties), mResource(std::move(resource)) {} +void SharedBufferMemory::DestroyImpl() { + ToBackend(GetDevice())->ReferenceUntilUnused(std::move(mResource)); +} + // static ResultOrError<Ref<SharedBufferMemory>> SharedBufferMemory::Create( Device* device, @@ -48,12 +59,11 @@ ComPtr<ID3D12Resource> d3d12Resource = descriptor->resource; - ID3D12Device* resourceDevice = nullptr; - d3d12Resource->GetDevice(__uuidof(*resourceDevice), reinterpret_cast<void**>(&resourceDevice)); - DAWN_INVALID_IF(resourceDevice != device->GetD3D12Device(), + ComPtr<ID3D12Device> resourceDevice; + d3d12Resource->GetDevice(__uuidof(resourceDevice), &resourceDevice); + DAWN_INVALID_IF(resourceDevice.Get() != device->GetD3D12Device(), "The D3D12 device of the resource and the D3D12 device of %s must be same.", device); - resourceDevice->Release(); D3D12_RESOURCE_DESC desc = d3d12Resource->GetDesc(); DAWN_INVALID_IF(desc.Dimension != D3D12_RESOURCE_DIMENSION_BUFFER, @@ -65,15 +75,31 @@ wgpu::BufferUsage usages = wgpu::BufferUsage::None; - if (desc.Flags & D3D12_RESOURCE_FLAG_ALLOW_UNORDERED_ACCESS) { - usages |= - wgpu::BufferUsage::Storage | wgpu::BufferUsage::CopySrc | wgpu::BufferUsage::CopyDst; - } else if (heapProperties.Type == D3D12_HEAP_TYPE_UPLOAD) { - usages |= wgpu::BufferUsage::MapWrite | wgpu::BufferUsage::CopySrc; - } else if (heapProperties.Type == D3D12_HEAP_TYPE_READBACK) { - usages |= wgpu::BufferUsage::MapRead | wgpu::BufferUsage::CopyDst; + switch (heapProperties.Type) { + case D3D12_HEAP_TYPE_UPLOAD: + usages |= wgpu::BufferUsage::MapWrite | wgpu::BufferUsage::CopySrc; + break; + case D3D12_HEAP_TYPE_READBACK: + usages |= wgpu::BufferUsage::MapRead | wgpu::BufferUsage::CopyDst; + break; + case D3D12_HEAP_TYPE_DEFAULT: + usages |= wgpu::BufferUsage::CopySrc | wgpu::BufferUsage::CopyDst | + wgpu::BufferUsage::Vertex | wgpu::BufferUsage::Index | + wgpu::BufferUsage::Indirect | wgpu::BufferUsage::QueryResolve; + if (desc.Flags & D3D12_RESOURCE_FLAG_ALLOW_UNORDERED_ACCESS) { + usages |= wgpu::BufferUsage::Storage; + } + if (IsAligned(desc.Width, D3D12_CONSTANT_BUFFER_DATA_PLACEMENT_ALIGNMENT)) { + usages |= wgpu::BufferUsage::Uniform; + } + break; + case D3D12_HEAP_TYPE_CUSTOM: + return DAWN_VALIDATION_ERROR( + "ID3D12Resources allocated on D3D12_HEAP_TYPE_CUSTOM heaps are not supported by " + "SharedBufferMemory."); + default: + DAWN_UNREACHABLE(); } - SharedBufferMemoryProperties properties; properties.size = desc.Width; properties.usage = usages; @@ -93,4 +119,42 @@ return mResource.Get(); } +MaybeError SharedBufferMemory::BeginAccessImpl( + BufferBase* buffer, + const UnpackedPtr<BeginAccessDescriptor>& descriptor) { + DAWN_TRY(descriptor.ValidateSubset<>()); + for (size_t i = 0; i < descriptor->fenceCount; ++i) { + SharedFenceBase* fence = descriptor->fences[i]; + + SharedFenceExportInfo exportInfo; + DAWN_TRY(fence->ExportInfo(&exportInfo)); + switch (exportInfo.type) { + case wgpu::SharedFenceType::DXGISharedHandle: + DAWN_INVALID_IF(!GetDevice()->HasFeature(Feature::SharedFenceDXGISharedHandle), + "Required feature (%s) is missing.", + wgpu::FeatureName::SharedFenceDXGISharedHandle); + break; + default: + return DAWN_VALIDATION_ERROR("Unsupported fence type %s.", exportInfo.type); + } + } + + return {}; +} + +ResultOrError<FenceAndSignalValue> SharedBufferMemory::EndAccessImpl( + BufferBase* buffer, + UnpackedPtr<EndAccessState>& state) { + DAWN_TRY(state.ValidateSubset<>()); + DAWN_INVALID_IF(!GetDevice()->HasFeature(Feature::SharedFenceDXGISharedHandle), + "Required feature (%s) is missing.", + wgpu::FeatureName::SharedFenceDXGISharedHandle); + + Ref<d3d::SharedFence> sharedFence; + DAWN_TRY_ASSIGN(sharedFence, ToBackend(GetDevice()->GetQueue())->GetOrCreateSharedFence()); + + return FenceAndSignalValue{std::move(sharedFence), + static_cast<uint64_t>(buffer->GetLastUsageSerial())}; +} + } // namespace dawn::native::d3d12
diff --git a/src/dawn/native/d3d12/SharedBufferMemoryD3D12.h b/src/dawn/native/d3d12/SharedBufferMemoryD3D12.h index 8f0fc38..53ccae8 100644 --- a/src/dawn/native/d3d12/SharedBufferMemoryD3D12.h +++ b/src/dawn/native/d3d12/SharedBufferMemoryD3D12.h
@@ -51,8 +51,14 @@ SharedBufferMemoryProperties properties, ComPtr<ID3D12Resource> resource); + void DestroyImpl() override; + ResultOrError<Ref<BufferBase>> CreateBufferImpl( const UnpackedPtr<BufferDescriptor>& descriptor) override; + MaybeError BeginAccessImpl(BufferBase* buffer, + const UnpackedPtr<BeginAccessDescriptor>& descriptor) override; + ResultOrError<FenceAndSignalValue> EndAccessImpl(BufferBase* buffer, + UnpackedPtr<EndAccessState>& state) override; ComPtr<ID3D12Resource> mResource; };
diff --git a/src/dawn/tests/white_box/SharedBufferMemoryTests.cpp b/src/dawn/tests/white_box/SharedBufferMemoryTests.cpp index f3ae18f..cb9c5ae 100644 --- a/src/dawn/tests/white_box/SharedBufferMemoryTests.cpp +++ b/src/dawn/tests/white_box/SharedBufferMemoryTests.cpp
@@ -26,9 +26,11 @@ // OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE. #include "dawn/tests/white_box/SharedBufferMemoryTests.h" + #include <gtest/gtest.h> #include <vector> #include "dawn/tests/DawnTest.h" +#include "dawn/utils/ComboRenderPipelineDescriptor.h" #include "dawn/utils/WGPUHelpers.h" namespace dawn { @@ -47,8 +49,74 @@ return features; } +void SharedBufferMemoryTests::MapAsyncAndWait(const wgpu::Buffer& buffer, + wgpu::MapMode mode, + uint32_t bufferSize) { + bool done = false; + buffer.MapAsync( + mode, 0, bufferSize, + [](WGPUBufferMapAsyncStatus status, void* userdata) { + ASSERT_EQ(WGPUBufferMapAsyncStatus_Success, status); + *static_cast<bool*>(userdata) = true; + }, + &done); + + while (!done) { + WaitABit(); + } +} + +wgpu::Texture Create2DTexture(wgpu::Device device, + uint32_t width, + uint32_t height, + wgpu::TextureFormat format, + wgpu::TextureUsage usage) { + wgpu::TextureDescriptor descriptor; + descriptor.dimension = wgpu::TextureDimension::e2D; + descriptor.size.width = width; + descriptor.size.height = height; + descriptor.size.depthOrArrayLayers = 1; + descriptor.sampleCount = 1; + descriptor.format = format; + descriptor.mipLevelCount = 1; + descriptor.usage = usage; + return device.CreateTexture(&descriptor); +} + +wgpu::SharedFence SharedBufferMemoryTestBackend::ImportFenceTo(const wgpu::Device& importingDevice, + const wgpu::SharedFence& fence) { + wgpu::SharedFenceExportInfo exportInfo; + fence.ExportInfo(&exportInfo); + + switch (exportInfo.type) { + case wgpu::SharedFenceType::DXGISharedHandle: { + wgpu::SharedFenceDXGISharedHandleExportInfo dxgiExportInfo; + exportInfo.nextInChain = &dxgiExportInfo; + fence.ExportInfo(&exportInfo); + + wgpu::SharedFenceDXGISharedHandleDescriptor dxgiDesc; + dxgiDesc.handle = dxgiExportInfo.handle; + + wgpu::SharedFenceDescriptor fenceDesc; + fenceDesc.nextInChain = &dxgiDesc; + return importingDevice.ImportSharedFence(&fenceDesc); + } + default: + DAWN_UNREACHABLE(); + } +} + namespace { +constexpr uint32_t kBufferData = 0x76543210; +constexpr uint32_t kBufferData2 = 0x01234567; +constexpr uint32_t kBufferSize = 4; +constexpr wgpu::BufferUsage kMapWriteUsages = + wgpu::BufferUsage::MapWrite | wgpu::BufferUsage::CopySrc; +constexpr wgpu::BufferUsage kMapReadUsages = + wgpu::BufferUsage::MapRead | wgpu::BufferUsage::CopyDst; +constexpr wgpu::BufferUsage kStorageUsages = + wgpu::BufferUsage::CopySrc | wgpu::BufferUsage::CopyDst | wgpu::BufferUsage::Storage; using ::testing::HasSubstr; // Test that it is an error to import shared buffer memory without a chained struct. @@ -70,7 +138,8 @@ // Test that SharedBufferMemory::IsDeviceLost() returns the expected value before and // after destroying the device. TEST_P(SharedBufferMemoryTests, CheckIsDeviceLostBeforeAndAfterDestroyingDevice) { - wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory(device); + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); EXPECT_FALSE(memory.IsDeviceLost()); device.Destroy(); @@ -80,7 +149,8 @@ // Test that SharedBufferMemory::IsDeviceLost() returns the expected value before and // after losing the device. TEST_P(SharedBufferMemoryTests, CheckIsDeviceLostBeforeAndAfterLosingDevice) { - wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory(device); + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); EXPECT_FALSE(memory.IsDeviceLost()); LoseDeviceForTesting(device); @@ -101,7 +171,8 @@ // Tests that creating SharedBufferMemory validates buffer size. TEST_P(SharedBufferMemoryTests, SizeValidation) { - wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory(device); + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); wgpu::SharedBufferMemoryProperties properties; memory.GetProperties(&properties); @@ -114,7 +185,8 @@ // Tests that creating SharedBufferMemory validates buffer usages. TEST_P(SharedBufferMemoryTests, UsageValidation) { - wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory(device); + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); wgpu::SharedBufferMemoryProperties properties; memory.GetProperties(&properties); @@ -136,6 +208,369 @@ } } +// Tests that creating SharedBufferMemory emits a specific error message if Uniform usage specified. +TEST_P(SharedBufferMemoryTests, UniformUsageValidation) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::SharedBufferMemoryProperties properties; + memory.GetProperties(&properties); + + wgpu::BufferDescriptor bufferDesc = {}; + bufferDesc.size = properties.size; + bufferDesc.usage = properties.usage | wgpu::BufferUsage::Uniform; + + ASSERT_DEVICE_ERROR_MSG(memory.CreateBuffer(&bufferDesc), HasSubstr("Uniform")); +} + +// Ensure that EndAccess cannot be called on a mapped or pending mapped buffer. +TEST_P(SharedBufferMemoryTests, CallEndAccessOnMappedBuffer) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer buffer = memory.CreateBuffer(); + wgpu::SharedBufferMemoryBeginAccessDescriptor desc; + memory.BeginAccess(buffer, &desc); + + bool done = false; + buffer.MapAsync( + wgpu::MapMode::Write, 0, sizeof(uint32_t), + [](WGPUBufferMapAsyncStatus status, void* userdata) { + ASSERT_EQ(WGPUBufferMapAsyncStatus_Success, status); + *static_cast<bool*>(userdata) = true; + }, + &done); + + // Calling EndAccess should generate an error even if the buffer has not completed being mapped. + wgpu::SharedBufferMemoryEndAccessState state; + ASSERT_DEVICE_ERROR(memory.EndAccess(buffer, &state)); + + while (!done) { + WaitABit(); + } + + // Calling EndAccess should generate an error after being mapped. + ASSERT_DEVICE_ERROR(memory.EndAccess(buffer, &state)); +} + +// Ensure no queue usage can occur before calling BeginAccess. +TEST_P(SharedBufferMemoryTests, EnsureNoQueueUsageBeforeBeginAccess) { + // We can't test this invalid scenario without validation. + DAWN_SUPPRESS_TEST_IF(HasToggleEnabled("skip_validation")); + + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer sharedBuffer = memory.CreateBuffer(); + + wgpu::BufferDescriptor descriptor; + descriptor.size = kBufferSize; + descriptor.usage = wgpu::BufferUsage::CopyDst; + wgpu::Buffer buffer = device.CreateBuffer(&descriptor); + + // Using the buffer in a submit without calling BeginAccess should cause an error. + wgpu::CommandEncoder encoder = device.CreateCommandEncoder(); + encoder.CopyBufferToBuffer(sharedBuffer, 0, buffer, 0, kBufferSize); + wgpu::CommandBuffer commandBuffer = encoder.Finish(); + ASSERT_DEVICE_ERROR(queue.Submit(1, &commandBuffer)); +} + +// Ensure mapping cannot occur before calling BeginAccess. +TEST_P(SharedBufferMemoryTests, EnsureNoMapUsageBeforeBeginAccess) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer sharedBuffer = memory.CreateBuffer(); + + // Mapping a buffer without calling BeginAccess should cause an error. + ASSERT_DEVICE_ERROR(sharedBuffer.MapAsync(wgpu::MapMode::Write, 0, 4, nullptr, nullptr)); +} + +// Ensure multiple buffers created from a SharedBufferMemory cannot be accessed simultaneously. +TEST_P(SharedBufferMemoryTests, EnsureNoSimultaneousAccess) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer sharedBuffer = memory.CreateBuffer(); + + wgpu::SharedBufferMemoryBeginAccessDescriptor desc; + memory.BeginAccess(sharedBuffer, &desc); + + wgpu::Buffer sharedBuffer2 = memory.CreateBuffer(); + ASSERT_DEVICE_ERROR(memory.BeginAccess(sharedBuffer2, &desc)); +} + +// Validate that calling EndAccess before BeginAccess produces an error. +TEST_P(SharedBufferMemoryTests, EnsureNoEndAccessBeforeBeginAccess) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer buffer = memory.CreateBuffer(); + + wgpu::SharedBufferMemoryEndAccessState state; + ASSERT_DEVICE_ERROR(memory.EndAccess(buffer, &state)); +} + +// Validate that calling EndAccess on a different buffer created from the same Shared is invalid. +TEST_P(SharedBufferMemoryTests, EndAccessOnDifferentBuffer) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer buffer = memory.CreateBuffer(); + wgpu::Buffer buffer2 = memory.CreateBuffer(); + + wgpu::SharedBufferMemoryBeginAccessDescriptor desc; + memory.BeginAccess(buffer, &desc); + + wgpu::SharedBufferMemoryEndAccessState state; + ASSERT_DEVICE_ERROR(memory.EndAccess(buffer2, &state)); + + // Ensure that calling EndAccess on the correct buffer still returns a fence. + memory.EndAccess(buffer, &state); + ASSERT_EQ(state.fenceCount, static_cast<size_t>(1)); + ASSERT_NE(state.fences[0], nullptr); +} + +// Validate that calling BeginAccess twice produces an error. +TEST_P(SharedBufferMemoryTests, EnsureNoDuplicateBeginAccessCalls) { + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer buffer = memory.CreateBuffer(); + + wgpu::SharedBufferMemoryBeginAccessDescriptor desc; + memory.BeginAccess(buffer, &desc); + ASSERT_DEVICE_ERROR(memory.BeginAccess(buffer, &desc)); +} + +// Ensure the BeginAccessDescriptor initialized parameter preserves or clears the buffer as +// necessary. +TEST_P(SharedBufferMemoryTests, BeginAccessInitialization) { + // Create a buffer with initialized data. + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer buffer = memory.CreateBuffer(); + + // Write data into the shared buffer. + wgpu::SharedBufferMemoryBeginAccessDescriptor beginAccessDesc; + beginAccessDesc.initialized = false; + memory.BeginAccess(buffer, &beginAccessDesc); + + MapAsyncAndWait(buffer, wgpu::MapMode::Write, kBufferSize); + + uint32_t* mappedData = static_cast<uint32_t*>(buffer.GetMappedRange(0, kBufferSize)); + memcpy(mappedData, &kBufferData, kBufferSize); + buffer.Unmap(); + + wgpu::SharedBufferMemoryEndAccessState endState; + memory.EndAccess(buffer, &endState); + + EXPECT_EQ(endState.initialized, true); + + // Pass fences from the previous operation to the next BeginAccessDescriptor to ensure + // operations are complete. + std::vector<wgpu::SharedFence> sharedFences(endState.fenceCount); + for (size_t j = 0; j < endState.fenceCount; ++j) { + sharedFences[j] = GetParam().mBackend->ImportFenceTo(device, endState.fences[j]); + } + beginAccessDesc.fenceCount = sharedFences.size(); + beginAccessDesc.fences = sharedFences.data(); + beginAccessDesc.signaledValues = endState.signaledValues; + + // Create a second buffer from the SharedBuffer memory, which will be marked as initialized in + // the BeginAccessDescriptor. This buffer should preserve the data from the previous copy. + wgpu::Buffer buffer2 = memory.CreateBuffer(); + beginAccessDesc.initialized = true; + memory.BeginAccess(buffer2, &beginAccessDesc); + // The buffer should contain the data from initialization. + EXPECT_BUFFER_U32_EQ(kBufferData, buffer2, 0); + memory.EndAccess(buffer2, &endState); + + // Pass fences from the previous operation to the next BeginAccessDescriptor to ensure + // operations are complete. + std::vector<wgpu::SharedFence> sharedFences2(endState.fenceCount); + for (size_t j = 0; j < endState.fenceCount; ++j) { + sharedFences2[j] = GetParam().mBackend->ImportFenceTo(device, endState.fences[j]); + } + beginAccessDesc.fenceCount = sharedFences2.size(); + beginAccessDesc.fences = sharedFences2.data(); + beginAccessDesc.signaledValues = endState.signaledValues; + + // Create another buffer from the SharedBufferMemory, but mark it uninitialized in the + // BeginAccessDescriptor. + wgpu::Buffer buffer3 = memory.CreateBuffer(); + beginAccessDesc.initialized = false; + memory.BeginAccess(buffer3, &beginAccessDesc); + // The buffer should be zero'd out because the BeginAccessDescriptor stated it was + // uninitialized. + EXPECT_BUFFER_U32_EQ(0, buffer3, 0); + memory.EndAccess(buffer3, &endState); +} + +// Tests that an unininitialized buffer that is not read or writt +TEST_P(SharedBufferMemoryTests, UninitializedBufferRemainsUninitialized) { + // Create a buffer with initialized data. + wgpu::SharedBufferMemory memory = + GetParam().mBackend->CreateSharedBufferMemory(device, kMapWriteUsages, kBufferSize); + wgpu::Buffer buffer = memory.CreateBuffer(); + + wgpu::SharedBufferMemoryBeginAccessDescriptor beginAccessDesc; + beginAccessDesc.initialized = false; + memory.BeginAccess(buffer, &beginAccessDesc); + wgpu::SharedBufferMemoryEndAccessState state; + memory.EndAccess(buffer, &state); + ASSERT_EQ(state.initialized, false); +} + +// Read and write a buffer with MapWrite and CopySrc usages. +TEST_P(SharedBufferMemoryTests, ReadWriteSharedMapWriteBuffer) { + // Create buffer buffer with initialized data. + wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory( + device, kMapWriteUsages, kBufferSize, kBufferData); + wgpu::Buffer buffer = memory.CreateBuffer(); + + // Begin access and check the contents within Dawn. + wgpu::SharedBufferMemoryBeginAccessDescriptor beginAccessDesc; + beginAccessDesc.initialized = true; + memory.BeginAccess(buffer, &beginAccessDesc); + EXPECT_BUFFER_U32_EQ(kBufferData, buffer, 0); + + MapAsyncAndWait(buffer, wgpu::MapMode::Write, kBufferSize); + + uint32_t* mappedData = static_cast<uint32_t*>(buffer.GetMappedRange(0, kBufferSize)); + memcpy(mappedData, &kBufferData2, kBufferSize); + buffer.Unmap(); + + EXPECT_BUFFER_U32_EQ(kBufferData2, buffer, 0); +} + +// Read and write a buffer with MapRead and CopyDst usages. +TEST_P(SharedBufferMemoryTests, ReadWriteSharedMapReadBuffer) { + // Create buffer buffer with initialized data. + wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory( + device, kMapReadUsages, kBufferSize, kBufferData); + wgpu::Buffer buffer = memory.CreateBuffer(); + + // Begin access and check the contents within Dawn. + wgpu::SharedBufferMemoryBeginAccessDescriptor beginAccessDesc; + beginAccessDesc.initialized = true; + memory.BeginAccess(buffer, &beginAccessDesc); + + MapAsyncAndWait(buffer, wgpu::MapMode::Read, kBufferSize); + + const uint32_t* mappedData = + static_cast<const uint32_t*>(buffer.GetConstMappedRange(0, kBufferSize)); + ASSERT_EQ(*mappedData, kBufferData); + + buffer.Unmap(); + + // Copy new data into the buffer from within Dawn and check the contents. + wgpu::Buffer dawnBuffer = + utils::CreateBufferFromData(device, &kBufferData2, kBufferSize, wgpu::BufferUsage::CopySrc); + wgpu::CommandEncoder encoder = device.CreateCommandEncoder(); + encoder.CopyBufferToBuffer(dawnBuffer, 0, buffer, 0, 4); + wgpu::CommandBuffer commandBuffer = encoder.Finish(); + queue.Submit(1, &commandBuffer); + + MapAsyncAndWait(buffer, wgpu::MapMode::Read, kBufferSize); + + mappedData = static_cast<const uint32_t*>(buffer.GetConstMappedRange(0, kBufferSize)); + ASSERT_EQ(*mappedData, kBufferData2); +} + +// Test ensures that a shader can read and write from a shared storage buffer. +TEST_P(SharedBufferMemoryTests, ReadWriteSharedStorageBuffer) { + wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory( + device, kStorageUsages, kBufferSize, kBufferData); + wgpu::Buffer buffer = memory.CreateBuffer(); + + // Begin access and check the contents within Dawn. + wgpu::SharedBufferMemoryBeginAccessDescriptor beginAccessDesc; + beginAccessDesc.initialized = true; + memory.BeginAccess(buffer, &beginAccessDesc); + + wgpu::ComputePipelineDescriptor pipelineDescriptor; + + // This compute shader reads from the shared storage buffer and increments it by one. + pipelineDescriptor.compute.module = utils::CreateShaderModule(device, R"( + struct OutputBuffer { + value : u32 + } + + @group(0) @binding(0) var<storage, read_write> outputBuffer : OutputBuffer; + + @compute @workgroup_size(1) fn main() { + outputBuffer.value = outputBuffer.value + 1u; + })"); + + wgpu::ComputePipeline pipeline = device.CreateComputePipeline(&pipelineDescriptor); + wgpu::BindGroup bindGroup = + utils::MakeBindGroup(device, pipeline.GetBindGroupLayout(0), {{0, buffer}}); + + wgpu::CommandBuffer commands; + wgpu::CommandEncoder encoder = device.CreateCommandEncoder(); + wgpu::ComputePassEncoder pass = encoder.BeginComputePass(); + pass.SetPipeline(pipeline); + pass.SetBindGroup(0, bindGroup); + pass.DispatchWorkgroups(1); + pass.End(); + commands = encoder.Finish(); + queue.Submit(1, &commands); + + // The storage buffer should have been incremented by one in the compute shader. + EXPECT_BUFFER_U32_EQ(kBufferData + 1, buffer, 0); +} + +TEST_P(SharedBufferMemoryTests, ImportExportSharedFences) { + wgpu::SharedBufferMemory memory = GetParam().mBackend->CreateSharedBufferMemory( + device, kStorageUsages, kBufferSize, kBufferData); + wgpu::Buffer buffer = memory.CreateBuffer(); + wgpu::SharedBufferMemoryEndAccessState endState; + + // Each loop checks the value of a storage buffer is correct and increments the value in a + // compute shader. Every loop exports a shared fence, which will be imported in the next loop. + for (int i = 0; i < 5; i++) { + // Begin access and check the contents within Dawn. + wgpu::SharedBufferMemoryBeginAccessDescriptor beginAccessDesc; + beginAccessDesc.initialized = true; + + // Get any fences from the previous loop's SharedBufferMemoryEndAccessState. + std::vector<wgpu::SharedFence> sharedFences(endState.fenceCount); + for (size_t j = 0; j < endState.fenceCount; ++j) { + sharedFences[j] = GetParam().mBackend->ImportFenceTo(device, endState.fences[j]); + } + beginAccessDesc.fenceCount = sharedFences.size(); + beginAccessDesc.fences = sharedFences.data(); + beginAccessDesc.signaledValues = endState.signaledValues; + memory.BeginAccess(buffer, &beginAccessDesc); + + // The storage buffer should be incremented by one per loop + EXPECT_BUFFER_U32_EQ(kBufferData + i, buffer, 0); + + wgpu::ComputePipelineDescriptor pipelineDescriptor; + + // This compute shader reads from the shared storage buffer and increments it by one. + pipelineDescriptor.compute.module = utils::CreateShaderModule(device, R"( + struct OutputBuffer { + value : u32 + } + + @group(0) @binding(0) var<storage, read_write> outputBuffer : OutputBuffer; + + @compute @workgroup_size(1) fn main() { + outputBuffer.value = outputBuffer.value + 1u; + })"); + + wgpu::ComputePipeline pipeline = device.CreateComputePipeline(&pipelineDescriptor); + wgpu::BindGroup bindGroup = + utils::MakeBindGroup(device, pipeline.GetBindGroupLayout(0), {{0, buffer}}); + + wgpu::CommandBuffer commands; + wgpu::CommandEncoder encoder = device.CreateCommandEncoder(); + wgpu::ComputePassEncoder pass = encoder.BeginComputePass(); + pass.SetPipeline(pipeline); + pass.SetBindGroup(0, bindGroup); + pass.DispatchWorkgroups(1); + pass.End(); + commands = encoder.Finish(); + queue.Submit(1, &commands); + + memory.EndAccess(buffer, &endState); + } +} + GTEST_ALLOW_UNINSTANTIATED_PARAMETERIZED_TEST(SharedBufferMemoryTests); } // anonymous namespace
diff --git a/src/dawn/tests/white_box/SharedBufferMemoryTests.h b/src/dawn/tests/white_box/SharedBufferMemoryTests.h index b1490af..d7e3b14 100644 --- a/src/dawn/tests/white_box/SharedBufferMemoryTests.h +++ b/src/dawn/tests/white_box/SharedBufferMemoryTests.h
@@ -47,7 +47,14 @@ virtual std::vector<wgpu::FeatureName> RequiredFeatures(const wgpu::Adapter& device) const = 0; // Create one basic shared buffer memory. It should support most operations. - virtual wgpu::SharedBufferMemory CreateSharedBufferMemory(const wgpu::Device& device) = 0; + virtual wgpu::SharedBufferMemory CreateSharedBufferMemory(const wgpu::Device& device, + wgpu::BufferUsage usages, + uint32_t bufferSize, + uint32_t data = 0) = 0; + + // Creates a SharedFence from a backend-specific fence type. + wgpu::SharedFence ImportFenceTo(const wgpu::Device& importingDevice, + const wgpu::SharedFence& fence); }; using Backend = SharedBufferMemoryTestBackend*; @@ -57,6 +64,9 @@ public: void SetUp() override; std::vector<wgpu::FeatureName> GetRequiredFeatures() override; + + protected: + void MapAsyncAndWait(const wgpu::Buffer& buffer, wgpu::MapMode mode, uint32_t bufferSize); }; } // namespace dawn
diff --git a/src/dawn/tests/white_box/SharedBufferMemoryTests_win.cpp b/src/dawn/tests/white_box/SharedBufferMemoryTests_win.cpp index 06f5049..f70374a 100644 --- a/src/dawn/tests/white_box/SharedBufferMemoryTests_win.cpp +++ b/src/dawn/tests/white_box/SharedBufferMemoryTests_win.cpp
@@ -36,8 +36,53 @@ namespace dawn { namespace { +constexpr uint32_t kBufferSize = 4; -constexpr uint32_t kBufferWidth = 32; +struct FenceInfo { + ComPtr<ID3D12Fence> fence; + uint64_t signaledValue; +}; + +void WriteD3D12UploadBuffer(ID3D12Resource* resource, uint32_t data) { + void* mappedBufferBegin; + D3D12_RANGE range; + range.Begin = 0; + range.End = kBufferSize; + resource->Map(0, &range, &mappedBufferBegin); + memcpy(mappedBufferBegin, &data, kBufferSize); + resource->Unmap(0, &range); +} + +void CopyD3D12Resource(ID3D12Device* device, ID3D12Resource* source, ID3D12Resource* destination) { + ComPtr<ID3D12CommandAllocator> commandAllocator; + device->CreateCommandAllocator(D3D12_COMMAND_LIST_TYPE_DIRECT, IID_PPV_ARGS(&commandAllocator)); + ComPtr<ID3D12CommandQueue> commandQueue; + D3D12_COMMAND_QUEUE_DESC queueDesc = {}; + queueDesc.Flags = D3D12_COMMAND_QUEUE_FLAG_NONE; + queueDesc.Type = D3D12_COMMAND_LIST_TYPE_DIRECT; + device->CreateCommandQueue(&queueDesc, IID_PPV_ARGS(&commandQueue)); + ComPtr<ID3D12GraphicsCommandList> commandList; + + device->CreateCommandList(0, D3D12_COMMAND_LIST_TYPE_DIRECT, commandAllocator.Get(), nullptr, + IID_PPV_ARGS(&commandList)); + + ID3D12CommandList* commandLists[] = {commandList.Get()}; + commandList->CopyResource(destination, source); + commandList->Close(); + + commandQueue->ExecuteCommandLists(_countof(commandLists), commandLists); + + ComPtr<ID3D12Fence> fence; + device->CreateFence(0, D3D12_FENCE_FLAG_SHARED, IID_PPV_ARGS(&fence)); + UINT64 signaledValue = 1; + commandQueue->Signal(fence.Get(), signaledValue); + + HANDLE fenceEvent = 0; + if (fence->GetCompletedValue() < signaledValue) { + fence->SetEventOnCompletion(signaledValue, fenceEvent); + WaitForSingleObject(fenceEvent, INFINITE); + } +} class Backend : public SharedBufferMemoryTestBackend { public: @@ -47,13 +92,50 @@ } std::vector<wgpu::FeatureName> RequiredFeatures(const wgpu::Adapter& adapter) const override { - return {wgpu::FeatureName::SharedBufferMemoryD3D12Resource}; + return {wgpu::FeatureName::SharedBufferMemoryD3D12Resource, + wgpu::FeatureName::SharedFenceDXGISharedHandle}; } - wgpu::SharedBufferMemory CreateSharedBufferMemory(const wgpu::Device& device) override { + wgpu::SharedBufferMemory CreateSharedBufferMemory(const wgpu::Device& device, + wgpu::BufferUsage usages, + uint32_t bufferSize, + uint32_t initializationData = 0) override { ComPtr<ID3D12Device> d3d12Device = CreateD3D12Device(device); + + D3D12_HEAP_TYPE d3d12HeapType; + + if (usages & wgpu::BufferUsage::MapWrite) { + d3d12HeapType = D3D12_HEAP_TYPE_UPLOAD; + } else if (usages & wgpu::BufferUsage::MapRead) { + d3d12HeapType = D3D12_HEAP_TYPE_READBACK; + } else { + d3d12HeapType = D3D12_HEAP_TYPE_DEFAULT; + } + + // To use a buffer with CreateConstantBufferView, it must be aligned to a constant. + if (usages & wgpu::BufferUsage::Uniform) { + bufferSize = Align(bufferSize, D3D12_CONSTANT_BUFFER_DATA_PLACEMENT_ALIGNMENT); + } + ComPtr<ID3D12Resource> d3d12Resource = - CreateD3D12Buffer(d3d12Device.Get(), D3D12_HEAP_TYPE_UPLOAD, D3D12_RESOURCE_FLAG_NONE); + CreateD3D12Buffer(d3d12Device.Get(), d3d12HeapType, bufferSize); + + if (initializationData) { + switch (d3d12HeapType) { + case D3D12_HEAP_TYPE_UPLOAD: + WriteD3D12UploadBuffer(d3d12Resource.Get(), initializationData); + break; + case D3D12_HEAP_TYPE_READBACK: + case D3D12_HEAP_TYPE_DEFAULT: { + ComPtr<ID3D12Resource> uploadBuffer = + CreateD3D12Buffer(d3d12Device.Get(), D3D12_HEAP_TYPE_UPLOAD, bufferSize); + WriteD3D12UploadBuffer(uploadBuffer.Get(), initializationData); + CopyD3D12Resource(d3d12Device.Get(), uploadBuffer.Get(), d3d12Resource.Get()); + } break; + default: + DAWN_UNREACHABLE(); + } + } wgpu::SharedBufferMemoryDescriptor desc; native::d3d12::SharedBufferMemoryD3D12ResourceDescriptor sharedD3d12ResourceDesc; @@ -62,16 +144,19 @@ return device.ImportSharedBufferMemory(&desc); } - private: - ComPtr<ID3D12Device> CreateD3D12Device(const wgpu::Device& device) { - ComPtr<IDXGIAdapter> dxgiAdapter = native::d3d::GetDXGIAdapter(device.GetAdapter().Get()); - DXGI_ADAPTER_DESC adapterDesc; - dxgiAdapter->GetDesc(&adapterDesc); - + ComPtr<ID3D12Device> CreateD3D12Device(const wgpu::Device& device, + bool createWarpDevice = false) { + ComPtr<IDXGIAdapter> dxgiAdapter = nullptr; ComPtr<IDXGIFactory4> dxgiFactory; CreateDXGIFactory2(0, IID_PPV_ARGS(&dxgiFactory)); - dxgiAdapter = nullptr; - dxgiFactory->EnumAdapterByLuid(adapterDesc.AdapterLuid, IID_PPV_ARGS(&dxgiAdapter)); + if (createWarpDevice) { + dxgiFactory->EnumWarpAdapter(IID_PPV_ARGS(&dxgiAdapter)); + } else { + dxgiAdapter = native::d3d::GetDXGIAdapter(device.GetAdapter().Get()); + DXGI_ADAPTER_DESC adapterDesc; + dxgiAdapter->GetDesc(&adapterDesc); + dxgiFactory->EnumAdapterByLuid(adapterDesc.AdapterLuid, IID_PPV_ARGS(&dxgiAdapter)); + } ComPtr<ID3D12Device> d3d12Device; @@ -83,10 +168,19 @@ ComPtr<ID3D12Resource> CreateD3D12Buffer(ID3D12Device* device, D3D12_HEAP_TYPE heapType, - D3D12_RESOURCE_FLAGS resourceFlags) { - D3D12_RESOURCE_STATES initialResourceState = D3D12_RESOURCE_STATE_COMMON; - if (heapType == D3D12_HEAP_TYPE_UPLOAD) { - initialResourceState = D3D12_RESOURCE_STATE_GENERIC_READ; + uint32_t bufferSize = kBufferSize) { + D3D12_RESOURCE_STATES initialResourceState; + D3D12_RESOURCE_FLAGS resourceFlags = D3D12_RESOURCE_FLAG_NONE; + switch (heapType) { + case D3D12_HEAP_TYPE_UPLOAD: + initialResourceState = D3D12_RESOURCE_STATE_GENERIC_READ; + break; + case D3D12_HEAP_TYPE_READBACK: + initialResourceState = D3D12_RESOURCE_STATE_COPY_DEST; + break; + default: + initialResourceState = D3D12_RESOURCE_STATE_COMMON; + resourceFlags = D3D12_RESOURCE_FLAG_ALLOW_UNORDERED_ACCESS; } D3D12_HEAP_PROPERTIES heapProperties = {heapType, D3D12_CPU_PAGE_PROPERTY_UNKNOWN, @@ -95,7 +189,7 @@ D3D12_RESOURCE_DESC descriptor; descriptor.Dimension = D3D12_RESOURCE_DIMENSION_BUFFER; descriptor.Alignment = 0; - descriptor.Width = kBufferWidth; + descriptor.Width = bufferSize; descriptor.Height = 1; descriptor.DepthOrArraySize = 1; descriptor.MipLevels = 1; @@ -116,26 +210,36 @@ Backend() {} }; -// TODO(dawn:2382): Add D3D12-specific tests for: -// - Test importing an {UPLOAD, READBACK, DEFAULT} buffer. -// - Test reading to an {UPLOAD, READBACK, DEFAULT} buffer. -// - Test Writing to an {UPLOAD, READBACK, DEFAULT} buffer -// - Ensure BeginAccess works with SharedFence. -// - Ensure EndAccess works with SharedFence. -// - Validate that importing a nullptr ID3D12Resource results in error. -// - Validate that importing an ID3D12Resource from another device results in error. -// - Check using a non-mappable buffers between multiple devices. -// - Check using the mappable buffers between multiple devices -// - Check validation that isInitialized must be true (for now). +// Ensure that importing a nullptr ID3D12Resource results in error. +TEST_P(SharedBufferMemoryTests, nullResourceFailure) { + native::d3d12::SharedBufferMemoryD3D12ResourceDescriptor sharedD3d12ResourceDesc; + sharedD3d12ResourceDesc.resource = nullptr; + wgpu::SharedBufferMemoryDescriptor desc; + desc.nextInChain = &sharedD3d12ResourceDesc; + ASSERT_DEVICE_ERROR(device.ImportSharedBufferMemory(&desc)); +} -// TODO(dawn:2382): Add backend-agnostic tests for: -// - Ensure that EndAccess cannot be called on a mapped buffer. -// - Ensure no operations {mapping, use on queue} can occur before calling BeginAccess. -// - Ensure multiple buffers created from a SharedBufferMemory cannot be accessed simultaneously. -// - Validate that calling EndAccess before BeginAccess produces an error. -// - Validate that calling BeginAccess twice produces an error. +// Validate that importing an ID3D12Resource across devices results in failure. This is tested by +// creating a resource with a WARP device and attempting to use it on a non-WARP device. +TEST_P(SharedBufferMemoryTests, CrossDeviceResourceImportFailure) { + DAWN_TEST_UNSUPPORTED_IF(IsWARP()); + ComPtr<ID3D12Device> warpDevice = + static_cast<Backend*>(GetParam().mBackend)->CreateD3D12Device(device, true); + ComPtr<ID3D12Resource> d3d12Resource = + static_cast<Backend*>(GetParam().mBackend) + ->CreateD3D12Buffer(warpDevice.Get(), D3D12_HEAP_TYPE_UPLOAD, D3D12_RESOURCE_FLAG_NONE); + wgpu::SharedBufferMemoryDescriptor desc; + native::d3d12::SharedBufferMemoryD3D12ResourceDescriptor sharedD3d12ResourceDesc; + sharedD3d12ResourceDesc.resource = d3d12Resource.Get(); + desc.nextInChain = &sharedD3d12ResourceDesc; -DAWN_INSTANTIATE_TEST_P(SharedBufferMemoryTests, {D3D12Backend()}, {Backend::GetInstance()}); + ASSERT_DEVICE_ERROR(device.ImportSharedBufferMemory(&desc)); +} + +DAWN_INSTANTIATE_PREFIXED_TEST_P(D3D12, + SharedBufferMemoryTests, + {D3D12Backend()}, + {Backend::GetInstance()}); } // anonymous namespace } // namespace dawn