encounter
/
dawn-cmake
mirror of https://github.com/encounter/dawn-cmake.git
1
0
Fork 0
Code Issues Packages Projects Releases Wiki Activity

Implement timestampWrites on D3D12 and Vulkan backends 

Metal implementation is more complex and has more issues, submit D3D12
and Vulkan frist and another CL for Metal.

Bug: dawn:1250
Change-Id: I718d323e01bb41b0209bfd1f1026faf64b4f1076
Reviewed-on: https://dawn-review.googlesource.com/c/dawn/+/97640
Reviewed-by: Austin Eng <enga@chromium.org>
Commit-Queue: Hao Li <hao.x.li@intel.com>
This commit is contained in:
Li Hao 2022-08-02 00:20:34 +00:00 committed by Dawn LUCI CQ
parent 9ec183ed9f
commit 9afd92be3f
13 changed files with 403 additions and 120 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

71
src/dawn/native/CommandEncoder.cpp
View File

@ -778,8 +778,6 @@ ComputePassEncoder* CommandEncoder::APIBeginComputePass(const ComputePassDescrip
Ref<ComputePassEncoder> CommandEncoder::BeginComputePass(const ComputePassDescriptor* descriptor) {
DeviceBase* device = GetDevice();
std::vector<TimestampWrite> timestampWritesAtBeginning;
std::vector<TimestampWrite> timestampWritesAtEnd;
bool success = mEncodingContext.TryEncode(
this,
[&](CommandAllocator* allocator) -> MaybeError {
@ -792,27 +790,27 @@ Ref<ComputePassEncoder> CommandEncoder::BeginComputePass(const ComputePassDescri
return {};
}
// Split the timestampWrites used in BeginComputePassCmd and EndComputePassCmd
// Record timestamp writes at the beginning and end of compute pass. The timestamp write
// at the end also be needed in BeginComputePassCmd because it's required by compute
// pass descriptor when beginning compute pass on Metal.
for (uint32_t i = 0; i < descriptor->timestampWriteCount; i++) {
QuerySetBase* querySet = descriptor->timestampWrites[i].querySet;
uint32_t queryIndex = descriptor->timestampWrites[i].queryIndex;
switch (descriptor->timestampWrites[i].location) {
case wgpu::ComputePassTimestampLocation::Beginning:
timestampWritesAtBeginning.push_back({querySet, queryIndex});
cmd->beginTimestamp.querySet = querySet;
cmd->beginTimestamp.queryIndex = queryIndex;
break;
case wgpu::ComputePassTimestampLocation::End:
timestampWritesAtEnd.push_back({querySet, queryIndex});
break;
default:
cmd->endTimestamp.querySet = querySet;
cmd->endTimestamp.queryIndex = queryIndex;
break;
}
TrackQueryAvailability(querySet, queryIndex);
}
cmd->timestampWrites = std::move(timestampWritesAtBeginning);
return {};
},
"encoding %s.BeginComputePass(%s).", this, descriptor);
@ -823,8 +821,8 @@ Ref<ComputePassEncoder> CommandEncoder::BeginComputePass(const ComputePassDescri
descriptor = &defaultDescriptor;
}
Ref<ComputePassEncoder> passEncoder = ComputePassEncoder::Create(
device, descriptor, this, &mEncodingContext, std::move(timestampWritesAtEnd));
Ref<ComputePassEncoder> passEncoder =
ComputePassEncoder::Create(device, descriptor, this, &mEncodingContext);
mEncodingContext.EnterPass(passEncoder.Get());
return passEncoder;
}
@ -846,8 +844,6 @@ Ref<RenderPassEncoder> CommandEncoder::BeginRenderPass(const RenderPassDescripto
bool depthReadOnly = false;
bool stencilReadOnly = false;
Ref<AttachmentState> attachmentState;
std::vector<TimestampWrite> timestampWritesAtBeginning;
std::vector<TimestampWrite> timestampWritesAtEnd;
bool success = mEncodingContext.TryEncode(
this,
[&](CommandAllocator* allocator) -> MaybeError {
@ -865,28 +861,6 @@ Ref<RenderPassEncoder> CommandEncoder::BeginRenderPass(const RenderPassDescripto
cmd->attachmentState = device->GetOrCreateAttachmentState(descriptor);
attachmentState = cmd->attachmentState;
// Split the timestampWrites used in BeginRenderPassCmd and EndRenderPassCmd
for (uint32_t i = 0; i < descriptor->timestampWriteCount; i++) {
QuerySetBase* querySet = descriptor->timestampWrites[i].querySet;
uint32_t queryIndex = descriptor->timestampWrites[i].queryIndex;
switch (descriptor->timestampWrites[i].location) {
case wgpu::RenderPassTimestampLocation::Beginning:
timestampWritesAtBeginning.push_back({querySet, queryIndex});
break;
case wgpu::RenderPassTimestampLocation::End:
timestampWritesAtEnd.push_back({querySet, queryIndex});
break;
default:
break;
}
TrackQueryAvailability(querySet, queryIndex);
// Track the query availability with true on render pass again for rewrite
// validation and query reset on Vulkan
usageTracker.TrackQueryAvailability(querySet, queryIndex);
}
for (ColorAttachmentIndex index :
IterateBitSet(cmd->attachmentState->GetColorAttachmentsMask())) {
uint8_t i = static_cast<uint8_t>(index);
@ -990,7 +964,29 @@ Ref<RenderPassEncoder> CommandEncoder::BeginRenderPass(const RenderPassDescripto
cmd->occlusionQuerySet = descriptor->occlusionQuerySet;
cmd->timestampWrites = std::move(timestampWritesAtBeginning);
// Record timestamp writes at the beginning and end of render pass. The timestamp write
// at the end also be needed in BeginComputePassCmd because it's required by render pass
// descriptor when beginning render pass on Metal.
for (uint32_t i = 0; i < descriptor->timestampWriteCount; i++) {
QuerySetBase* querySet = descriptor->timestampWrites[i].querySet;
uint32_t queryIndex = descriptor->timestampWrites[i].queryIndex;
switch (descriptor->timestampWrites[i].location) {
case wgpu::RenderPassTimestampLocation::Beginning:
cmd->beginTimestamp.querySet = querySet;
cmd->beginTimestamp.queryIndex = queryIndex;
break;
case wgpu::RenderPassTimestampLocation::End:
cmd->endTimestamp.querySet = querySet;
cmd->endTimestamp.queryIndex = queryIndex;
break;
}
TrackQueryAvailability(querySet, queryIndex);
// Track the query availability with true on render pass again for rewrite
// validation and query reset on Vulkan
usageTracker.TrackQueryAvailability(querySet, queryIndex);
}
return {};
},
@ -999,8 +995,7 @@ Ref<RenderPassEncoder> CommandEncoder::BeginRenderPass(const RenderPassDescripto
if (success) {
Ref<RenderPassEncoder> passEncoder = RenderPassEncoder::Create(
device, descriptor, this, &mEncodingContext, std::move(usageTracker),
std::move(attachmentState), std::move(timestampWritesAtEnd), width, height,
depthReadOnly, stencilReadOnly);
std::move(attachmentState), width, height, depthReadOnly, stencilReadOnly);
mEncodingContext.EnterPass(passEncoder.Get());
return passEncoder;
}

4
src/dawn/native/Commands.cpp
View File

@ -361,9 +361,7 @@ void SkipCommand(CommandIterator* commands, Command type) {
}
}
TimestampWrite::TimestampWrite(const Ref<QuerySetBase>& set, uint32_t idx)
: querySet(set), queryIndex(idx) {}
TimestampWrite::TimestampWrite(TimestampWrite&&) = default;
TimestampWrite::TimestampWrite() = default;
TimestampWrite::~TimestampWrite() = default;
BeginComputePassCmd::BeginComputePassCmd() = default;

13
src/dawn/native/Commands.h
View File

@ -70,8 +70,7 @@ enum class Command {
};
struct TimestampWrite {
TimestampWrite(const Ref<QuerySetBase>& set, uint32_t idx);
TimestampWrite(TimestampWrite&&);
TimestampWrite();
~TimestampWrite();
Ref<QuerySetBase> querySet;
@ -82,7 +81,8 @@ struct BeginComputePassCmd {
BeginComputePassCmd();
~BeginComputePassCmd();
std::vector<TimestampWrite> timestampWrites;
TimestampWrite beginTimestamp;
TimestampWrite endTimestamp;
};
struct BeginOcclusionQueryCmd {
@ -133,7 +133,8 @@ struct BeginRenderPassCmd {
uint32_t height;
Ref<QuerySetBase> occlusionQuerySet;
std::vector<TimestampWrite> timestampWrites;
TimestampWrite beginTimestamp;
TimestampWrite endTimestamp;
};
struct BufferCopy {
@ -228,8 +229,6 @@ struct DrawIndexedIndirectCmd : DrawIndirectCmd {};
struct EndComputePassCmd {
EndComputePassCmd();
~EndComputePassCmd();
std::vector<TimestampWrite> timestampWrites;
};
struct EndOcclusionQueryCmd {
@ -243,8 +242,6 @@ struct EndOcclusionQueryCmd {
struct EndRenderPassCmd {
EndRenderPassCmd();
~EndRenderPassCmd();
std::vector<TimestampWrite> timestampWrites;
};
struct ExecuteBundlesCmd {

19
src/dawn/native/ComputePassEncoder.cpp
View File

@ -111,23 +111,18 @@ ResultOrError<ComputePipelineBase*> GetOrCreateIndirectDispatchValidationPipelin
ComputePassEncoder::ComputePassEncoder(DeviceBase* device,
const ComputePassDescriptor* descriptor,
CommandEncoder* commandEncoder,
EncodingContext* encodingContext,
std::vector<TimestampWrite> timestampWritesAtEnd)
EncodingContext* encodingContext)
: ProgrammableEncoder(device, descriptor->label, encodingContext),
mCommandEncoder(commandEncoder),
mTimestampWritesAtEnd(std::move(timestampWritesAtEnd)) {
mCommandEncoder(commandEncoder) {
TrackInDevice();
}
// static
Ref<ComputePassEncoder> ComputePassEncoder::Create(
DeviceBase* device,
Ref<ComputePassEncoder> ComputePassEncoder::Create(DeviceBase* device,
const ComputePassDescriptor* descriptor,
CommandEncoder* commandEncoder,
EncodingContext* encodingContext,
std::vector<TimestampWrite> timestampWritesAtEnd) {
return AcquireRef(new ComputePassEncoder(device, descriptor, commandEncoder, encodingContext,
std::move(timestampWritesAtEnd)));
EncodingContext* encodingContext) {
return AcquireRef(new ComputePassEncoder(device, descriptor, commandEncoder, encodingContext));
}
ComputePassEncoder::ComputePassEncoder(DeviceBase* device,
@ -162,11 +157,7 @@ void ComputePassEncoder::APIEnd() {
DAWN_TRY(ValidateProgrammableEncoderEnd());
}
EndComputePassCmd* cmd =
allocator->Allocate<EndComputePassCmd>(Command::EndComputePass);
// The query availability has already been updated at the beginning of compute
// pass, and no need to do update here.
cmd->timestampWrites = std::move(mTimestampWritesAtEnd);
return {};
},

8
src/dawn/native/ComputePassEncoder.h
View File

@ -33,8 +33,7 @@ class ComputePassEncoder final : public ProgrammableEncoder {
static Ref<ComputePassEncoder> Create(DeviceBase* device,
const ComputePassDescriptor* descriptor,
CommandEncoder* commandEncoder,
EncodingContext* encodingContext,
std::vector<TimestampWrite> timestampWritesAtEnd);
EncodingContext* encodingContext);
static Ref<ComputePassEncoder> MakeError(DeviceBase* device,
CommandEncoder* commandEncoder,
EncodingContext* encodingContext);
@ -72,8 +71,7 @@ class ComputePassEncoder final : public ProgrammableEncoder {
ComputePassEncoder(DeviceBase* device,
const ComputePassDescriptor* descriptor,
CommandEncoder* commandEncoder,
EncodingContext* encodingContext,
std::vector<TimestampWrite> timestampWritesAtEnd);
EncodingContext* encodingContext);
ComputePassEncoder(DeviceBase* device,
CommandEncoder* commandEncoder,
EncodingContext* encodingContext,
@ -98,8 +96,6 @@ class ComputePassEncoder final : public ProgrammableEncoder {
// For render and compute passes, the encoding context is borrowed from the command encoder.
// Keep a reference to the encoder to make sure the context isn't freed.
Ref<CommandEncoder> mCommandEncoder;
std::vector<TimestampWrite> mTimestampWritesAtEnd;
};
} // namespace dawn::native

14
src/dawn/native/RenderPassEncoder.cpp
View File

@ -56,7 +56,6 @@ RenderPassEncoder::RenderPassEncoder(DeviceBase* device,
EncodingContext* encodingContext,
RenderPassResourceUsageTracker usageTracker,
Ref<AttachmentState> attachmentState,
std::vector<TimestampWrite> timestampWritesAtEnd,
uint32_t renderTargetWidth,
uint32_t renderTargetHeight,
bool depthReadOnly,
@ -70,8 +69,7 @@ RenderPassEncoder::RenderPassEncoder(DeviceBase* device,
mCommandEncoder(commandEncoder),
mRenderTargetWidth(renderTargetWidth),
mRenderTargetHeight(renderTargetHeight),
mOcclusionQuerySet(descriptor->occlusionQuerySet),
mTimestampWritesAtEnd(std::move(timestampWritesAtEnd)) {
mOcclusionQuerySet(descriptor->occlusionQuerySet) {
mUsageTracker = std::move(usageTracker);
const RenderPassDescriptorMaxDrawCount* maxDrawCountInfo = nullptr;
FindInChain(descriptor->nextInChain, &maxDrawCountInfo);
@ -88,15 +86,14 @@ Ref<RenderPassEncoder> RenderPassEncoder::Create(DeviceBase* device,
EncodingContext* encodingContext,
RenderPassResourceUsageTracker usageTracker,
Ref<AttachmentState> attachmentState,
std::vector<TimestampWrite> timestampWritesAtEnd,
uint32_t renderTargetWidth,
uint32_t renderTargetHeight,
bool depthReadOnly,
bool stencilReadOnly) {
return AcquireRef(new RenderPassEncoder(device, descriptor, commandEncoder, encodingContext,
std::move(usageTracker), std::move(attachmentState),
std::move(timestampWritesAtEnd), renderTargetWidth,
renderTargetHeight, depthReadOnly, stencilReadOnly));
renderTargetWidth, renderTargetHeight, depthReadOnly,
stencilReadOnly));
}
RenderPassEncoder::RenderPassEncoder(DeviceBase* device,
@ -152,10 +149,7 @@ void RenderPassEncoder::APIEnd() {
mDrawCount, this, mMaxDrawCount);
}
EndRenderPassCmd* cmd = allocator->Allocate<EndRenderPassCmd>(Command::EndRenderPass);
// The query availability has already been updated at the beginning of render
// pass, and no need to do update here.
cmd->timestampWrites = std::move(mTimestampWritesAtEnd);
allocator->Allocate<EndRenderPassCmd>(Command::EndRenderPass);
DAWN_TRY(mEncodingContext->ExitRenderPass(this, std::move(mUsageTracker),
mCommandEncoder.Get(),

4
src/dawn/native/RenderPassEncoder.h
View File

@ -33,7 +33,6 @@ class RenderPassEncoder final : public RenderEncoderBase {
EncodingContext* encodingContext,
RenderPassResourceUsageTracker usageTracker,
Ref<AttachmentState> attachmentState,
std::vector<TimestampWrite> timestampWritesAtEnd,
uint32_t renderTargetWidth,
uint32_t renderTargetHeight,
bool depthReadOnly,
@ -70,7 +69,6 @@ class RenderPassEncoder final : public RenderEncoderBase {
EncodingContext* encodingContext,
RenderPassResourceUsageTracker usageTracker,
Ref<AttachmentState> attachmentState,
std::vector<TimestampWrite> timestampWritesAtEnd,
uint32_t renderTargetWidth,
uint32_t renderTargetHeight,
bool depthReadOnly,
@ -99,8 +97,6 @@ class RenderPassEncoder final : public RenderEncoderBase {
// This is the hardcoded value in the WebGPU spec.
uint64_t mMaxDrawCount = 50000000;
std::vector<TimestampWrite> mTimestampWritesAtEnd;
};
} // namespace dawn::native

47
src/dawn/native/d3d12/CommandBufferD3D12.cpp
View File

@ -90,10 +90,12 @@ bool CanUseCopyResource(const TextureCopy& src, const TextureCopy& dst, const Ex
copySize.depthOrArrayLayers == srcSize.depthOrArrayLayers;
}
void RecordWriteTimestampCmd(ID3D12GraphicsCommandList* commandList, WriteTimestampCmd* cmd) {
QuerySet* querySet = ToBackend(cmd->querySet.Get());
ASSERT(D3D12QueryType(querySet->GetQueryType()) == D3D12_QUERY_TYPE_TIMESTAMP);
commandList->EndQuery(querySet->GetQueryHeap(), D3D12_QUERY_TYPE_TIMESTAMP, cmd->queryIndex);
void RecordWriteTimestampCmd(ID3D12GraphicsCommandList* commandList,
QuerySetBase* querySet,
uint32_t queryIndex) {
ASSERT(D3D12QueryType(ToBackend(querySet)->GetQueryType()) == D3D12_QUERY_TYPE_TIMESTAMP);
commandList->EndQuery(ToBackend(querySet)->GetQueryHeap(), D3D12_QUERY_TYPE_TIMESTAMP,
queryIndex);
}
void RecordResolveQuerySetCmd(ID3D12GraphicsCommandList* commandList,
@ -653,11 +655,12 @@ MaybeError CommandBuffer::RecordCommands(CommandRecordingContext* commandContext
while (mCommands.NextCommandId(&type)) {
switch (type) {
case Command::BeginComputePass: {
mCommands.NextCommand<BeginComputePassCmd>();
BeginComputePassCmd* cmd = mCommands.NextCommand<BeginComputePassCmd>();
bindingTracker.SetInComputePass(true);
DAWN_TRY(
RecordComputePass(commandContext, &bindingTracker,
RecordComputePass(commandContext, &bindingTracker, cmd,
GetResourceUsages().computePasses[nextComputePassNumber]));
nextComputePassNumber++;
@ -942,7 +945,7 @@ MaybeError CommandBuffer::RecordCommands(CommandRecordingContext* commandContext
case Command::WriteTimestamp: {
WriteTimestampCmd* cmd = mCommands.NextCommand<WriteTimestampCmd>();
RecordWriteTimestampCmd(commandList, cmd);
RecordWriteTimestampCmd(commandList, cmd->querySet.Get(), cmd->queryIndex);
break;
}
@ -1023,10 +1026,17 @@ MaybeError CommandBuffer::RecordCommands(CommandRecordingContext* commandContext
MaybeError CommandBuffer::RecordComputePass(CommandRecordingContext* commandContext,
BindGroupStateTracker* bindingTracker,
BeginComputePassCmd* computePass,
const ComputePassResourceUsage& resourceUsages) {
uint64_t currentDispatch = 0;
ID3D12GraphicsCommandList* commandList = commandContext->GetCommandList();
// Write timestamp at the beginning of compute pass if it's set.
if (computePass->beginTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(commandList, computePass->beginTimestamp.querySet.Get(),
computePass->beginTimestamp.queryIndex);
}
Command type;
ComputePipeline* lastPipeline = nullptr;
while (mCommands.NextCommandId(&type)) {
@ -1068,6 +1078,12 @@ MaybeError CommandBuffer::RecordComputePass(CommandRecordingContext* commandCont
case Command::EndComputePass: {
mCommands.NextCommand<EndComputePassCmd>();
// Write timestamp at the end of compute pass if it's set.
if (computePass->endTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(commandList, computePass->endTimestamp.querySet.Get(),
computePass->endTimestamp.queryIndex);
}
return {};
}
@ -1136,7 +1152,7 @@ MaybeError CommandBuffer::RecordComputePass(CommandRecordingContext* commandCont
case Command::WriteTimestamp: {
WriteTimestampCmd* cmd = mCommands.NextCommand<WriteTimestampCmd>();
RecordWriteTimestampCmd(commandList, cmd);
RecordWriteTimestampCmd(commandList, cmd->querySet.Get(), cmd->queryIndex);
break;
}
@ -1339,6 +1355,12 @@ MaybeError CommandBuffer::RecordRenderPass(CommandRecordingContext* commandConte
ID3D12GraphicsCommandList* commandList = commandContext->GetCommandList();
// Write timestamp at the beginning of render pass if it's set.
if (renderPass->beginTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(commandList, renderPass->beginTimestamp.querySet.Get(),
renderPass->beginTimestamp.queryIndex);
}
// Set up default dynamic state
{
uint32_t width = renderPass->width;
@ -1511,6 +1533,13 @@ MaybeError CommandBuffer::RecordRenderPass(CommandRecordingContext* commandConte
switch (type) {
case Command::EndRenderPass: {
mCommands.NextCommand<EndRenderPassCmd>();
// Write timestamp at the end of render pass if it's set.
if (renderPass->endTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(commandList, renderPass->endTimestamp.querySet.Get(),
renderPass->endTimestamp.queryIndex);
}
if (useRenderPass) {
commandContext->GetCommandList4()->EndRenderPass();
} else if (renderPass->attachmentState->GetSampleCount() > 1) {
@ -1596,7 +1625,7 @@ MaybeError CommandBuffer::RecordRenderPass(CommandRecordingContext* commandConte
case Command::WriteTimestamp: {
WriteTimestampCmd* cmd = mCommands.NextCommand<WriteTimestampCmd>();
RecordWriteTimestampCmd(commandList, cmd);
RecordWriteTimestampCmd(commandList, cmd->querySet.Get(), cmd->queryIndex);
break;
}

2
src/dawn/native/d3d12/CommandBufferD3D12.h
View File

@ -19,6 +19,7 @@
#include "dawn/native/Error.h"
namespace dawn::native {
struct BeginComputePassCmd;
struct BeginRenderPassCmd;
} // namespace dawn::native
@ -40,6 +41,7 @@ class CommandBuffer final : public CommandBufferBase {
MaybeError RecordComputePass(CommandRecordingContext* commandContext,
BindGroupStateTracker* bindingTracker,
BeginComputePassCmd* computePass,
const ComputePassResourceUsage& resourceUsages);
MaybeError RecordRenderPass(CommandRecordingContext* commandContext,
BindGroupStateTracker* bindingTracker,

12
src/dawn/native/metal/CommandBufferMTL.mm
View File

@ -610,7 +610,12 @@ MaybeError CommandBuffer::FillCommands(CommandRecordingContext* commandContext)
while (mCommands.NextCommandId(&type)) {
switch (type) {
case Command::BeginComputePass: {
mCommands.NextCommand<BeginComputePassCmd>();
BeginComputePassCmd* cmd = mCommands.NextCommand<BeginComputePassCmd>();
if (cmd->beginTimestamp.querySet.Get() != nullptr ||
cmd->endTimestamp.querySet.Get() != nullptr) {
return DAWN_UNIMPLEMENTED_ERROR("timestampWrites unimplemented.");
}
for (const SyncScopeResourceUsage& scope :
GetResourceUsages().computePasses[nextComputePassNumber].dispatchUsages) {
@ -627,6 +632,11 @@ MaybeError CommandBuffer::FillCommands(CommandRecordingContext* commandContext)
case Command::BeginRenderPass: {
BeginRenderPassCmd* cmd = mCommands.NextCommand<BeginRenderPassCmd>();
if (cmd->beginTimestamp.querySet.Get() != nullptr ||
cmd->endTimestamp.querySet.Get() != nullptr) {
return DAWN_UNIMPLEMENTED_ERROR("timestampWrites unimplemented.");
}
LazyClearSyncScope(GetResourceUsages().renderPasses[nextRenderPassNumber],
commandContext);
commandContext->EndBlit();

68
src/dawn/native/vulkan/CommandBufferVk.cpp
View File

@ -377,12 +377,19 @@ void ResetUsedQuerySetsOnRenderPass(Device* device,
void RecordWriteTimestampCmd(CommandRecordingContext* recordingContext,
Device* device,
WriteTimestampCmd* cmd) {
QuerySetBase* querySet,
uint32_t queryIndex,
bool isRenderPass) {
VkCommandBuffer commands = recordingContext->commandBuffer;
QuerySet* querySet = ToBackend(cmd->querySet.Get());
// The queries must be reset between uses, and the reset command cannot be called in render
// pass.
if (!isRenderPass) {
device->fn.CmdResetQueryPool(commands, ToBackend(querySet)->GetHandle(), queryIndex, 1);
}
device->fn.CmdWriteTimestamp(commands, VK_PIPELINE_STAGE_ALL_COMMANDS_BIT,
querySet->GetHandle(), cmd->queryIndex);
ToBackend(querySet)->GetHandle(), queryIndex);
}
void RecordResolveQuerySetCmd(VkCommandBuffer commands,
@ -735,10 +742,11 @@ MaybeError CommandBuffer::RecordCommands(CommandRecordingContext* recordingConte
}
case Command::BeginComputePass: {
mCommands.NextCommand<BeginComputePassCmd>();
BeginComputePassCmd* cmd = mCommands.NextCommand<BeginComputePassCmd>();
DAWN_TRY(RecordComputePass(
recordingContext, GetResourceUsages().computePasses[nextComputePassNumber]));
DAWN_TRY(
RecordComputePass(recordingContext, cmd,
GetResourceUsages().computePasses[nextComputePassNumber]));
nextComputePassNumber++;
break;
@ -777,11 +785,8 @@ MaybeError CommandBuffer::RecordCommands(CommandRecordingContext* recordingConte
case Command::WriteTimestamp: {
WriteTimestampCmd* cmd = mCommands.NextCommand<WriteTimestampCmd>();
// The query must be reset between uses.
device->fn.CmdResetQueryPool(commands, ToBackend(cmd->querySet)->GetHandle(),
cmd->queryIndex, 1);
RecordWriteTimestampCmd(recordingContext, device, cmd);
RecordWriteTimestampCmd(recordingContext, device, cmd->querySet.Get(),
cmd->queryIndex, false);
break;
}
@ -878,8 +883,17 @@ MaybeError CommandBuffer::RecordCommands(CommandRecordingContext* recordingConte
}
MaybeError CommandBuffer::RecordComputePass(CommandRecordingContext* recordingContext,
BeginComputePassCmd* computePassCmd,
const ComputePassResourceUsage& resourceUsages) {
Device* device = ToBackend(GetDevice());
// Write timestamp at the beginning of compute pass if it's set
if (computePassCmd->beginTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(recordingContext, device,
computePassCmd->beginTimestamp.querySet.Get(),
computePassCmd->beginTimestamp.queryIndex, false);
}
VkCommandBuffer commands = recordingContext->commandBuffer;
uint64_t currentDispatch = 0;
@ -890,6 +904,13 @@ MaybeError CommandBuffer::RecordComputePass(CommandRecordingContext* recordingCo
switch (type) {
case Command::EndComputePass: {
mCommands.NextCommand<EndComputePassCmd>();
// Write timestamp at the end of compute pass if it's set.
if (computePassCmd->endTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(recordingContext, device,
computePassCmd->endTimestamp.querySet.Get(),
computePassCmd->endTimestamp.queryIndex, false);
}
return {};
}
@ -996,11 +1017,8 @@ MaybeError CommandBuffer::RecordComputePass(CommandRecordingContext* recordingCo
case Command::WriteTimestamp: {
WriteTimestampCmd* cmd = mCommands.NextCommand<WriteTimestampCmd>();
// The query must be reset between uses.
device->fn.CmdResetQueryPool(commands, ToBackend(cmd->querySet)->GetHandle(),
cmd->queryIndex, 1);
RecordWriteTimestampCmd(recordingContext, device, cmd);
RecordWriteTimestampCmd(recordingContext, device, cmd->querySet.Get(),
cmd->queryIndex, false);
break;
}
@ -1020,6 +1038,13 @@ MaybeError CommandBuffer::RecordRenderPass(CommandRecordingContext* recordingCon
DAWN_TRY(RecordBeginRenderPass(recordingContext, device, renderPassCmd));
// Write timestamp at the beginning of render pass if it's set.
if (renderPassCmd->beginTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(recordingContext, device,
renderPassCmd->beginTimestamp.querySet.Get(),
renderPassCmd->beginTimestamp.queryIndex, true);
}
// Set the default value for the dynamic state
{
device->fn.CmdSetLineWidth(commands, 1.0f);
@ -1203,6 +1228,14 @@ MaybeError CommandBuffer::RecordRenderPass(CommandRecordingContext* recordingCon
switch (type) {
case Command::EndRenderPass: {
mCommands.NextCommand<EndRenderPassCmd>();
// Write timestamp at the end of render pass if it's set.
if (renderPassCmd->endTimestamp.querySet.Get() != nullptr) {
RecordWriteTimestampCmd(recordingContext, device,
renderPassCmd->endTimestamp.querySet.Get(),
renderPassCmd->endTimestamp.queryIndex, true);
}
device->fn.CmdEndRenderPass(commands);
return {};
}
@ -1290,7 +1323,8 @@ MaybeError CommandBuffer::RecordRenderPass(CommandRecordingContext* recordingCon
case Command::WriteTimestamp: {
WriteTimestampCmd* cmd = mCommands.NextCommand<WriteTimestampCmd>();
RecordWriteTimestampCmd(recordingContext, device, cmd);
RecordWriteTimestampCmd(recordingContext, device, cmd->querySet.Get(),
cmd->queryIndex, true);
break;
}

2
src/dawn/native/vulkan/CommandBufferVk.h
View File

@ -21,6 +21,7 @@
#include "dawn/common/vulkan_platform.h"
namespace dawn::native {
struct BeginComputePassCmd;
struct BeginRenderPassCmd;
struct TextureCopy;
} // namespace dawn::native
@ -41,6 +42,7 @@ class CommandBuffer final : public CommandBufferBase {
CommandBuffer(CommandEncoder* encoder, const CommandBufferDescriptor* descriptor);
MaybeError RecordComputePass(CommandRecordingContext* recordingContext,
BeginComputePassCmd* computePass,
const ComputePassResourceUsage& resourceUsages);
MaybeError RecordRenderPass(CommandRecordingContext* recordingContext,
BeginRenderPassCmd* renderPass);

253
src/dawn/tests/end2end/QueryTests.cpp
View File

@ -302,7 +302,7 @@ TEST_P(OcclusionQueryTests, Rewrite) {
// the query resetting at the start of render passes on Vulkan backend.
TEST_P(OcclusionQueryTests, ResolveSparseQueries) {
// TODO(hao.x.li@intel.com): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it util
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it until
// the issue is fixed.
DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsIntel());
@ -523,6 +523,17 @@ class TimestampQueryTests : public QueryTests {
// Skip all tests if timestamp feature is not supported
DAWN_TEST_UNSUPPORTED_IF(!SupportsFeatures({wgpu::FeatureName::TimestampQuery}));
// Create basic compute pipeline
wgpu::ShaderModule module = utils::CreateShaderModule(device, R"(
@compute @workgroup_size(1)
fn main() {
})");
wgpu::ComputePipelineDescriptor csDesc;
csDesc.compute.module = module;
csDesc.compute.entryPoint = "main";
computePipeline = device.CreateComputePipeline(&csDesc);
}
std::vector<wgpu::FeatureName> GetRequiredFeatures() override {
@ -539,6 +550,113 @@ class TimestampQueryTests : public QueryTests {
descriptor.type = wgpu::QueryType::Timestamp;
return device.CreateQuerySet(&descriptor);
}
void TestTimestampWritesOnComputePass(
const std::vector<wgpu::ComputePassTimestampWrite>& timestampWrites,
const std::vector<wgpu::ComputePassTimestampWrite>& timestampWritesOnAnotherPass = {}) {
size_t queryCount = timestampWrites.size() + timestampWritesOnAnotherPass.size();
// The destination buffer offset must be a multiple of 256.
wgpu::Buffer destination =
CreateResolveBuffer(queryCount * kMinDestinationOffset + sizeof(uint64_t));
wgpu::ComputePassDescriptor descriptor;
descriptor.timestampWriteCount = timestampWrites.size();
descriptor.timestampWrites = timestampWrites.data();
wgpu::CommandEncoder encoder = device.CreateCommandEncoder();
wgpu::ComputePassEncoder pass = encoder.BeginComputePass(&descriptor);
pass.SetPipeline(computePipeline);
pass.DispatchWorkgroups(1, 1, 1);
pass.End();
// Resolve queries one by one because the query set at the beginning of pass may be
// different with the one at the end of pass.
for (size_t i = 0; i < timestampWrites.size(); i++) {
encoder.ResolveQuerySet(timestampWrites[i].querySet, timestampWrites[i].queryIndex, 1,
destination, i * kMinDestinationOffset);
}
// Begin another compute pass if the timestampWritesOnAnotherPass is set.
if (!timestampWritesOnAnotherPass.empty()) {
wgpu::ComputePassDescriptor descriptor2;
descriptor2.timestampWriteCount = timestampWritesOnAnotherPass.size();
descriptor2.timestampWrites = timestampWritesOnAnotherPass.data();
wgpu::ComputePassEncoder pass2 = encoder.BeginComputePass(&descriptor2);
pass2.SetPipeline(computePipeline);
pass2.DispatchWorkgroups(1, 1, 1);
pass2.End();
for (size_t i = 0; i < timestampWritesOnAnotherPass.size(); i++) {
// Resolve queries one by one because the query set at the beginning of pass may be
// different with the one at the end of pass.
encoder.ResolveQuerySet(timestampWritesOnAnotherPass[i].querySet,
timestampWritesOnAnotherPass[i].queryIndex, 1, destination,
(timestampWrites.size() + i) * kMinDestinationOffset);
}
}
wgpu::CommandBuffer commands = encoder.Finish();
queue.Submit(1, &commands);
for (size_t i = 0; i < queryCount; i++) {
EXPECT_BUFFER(destination, i * kMinDestinationOffset, sizeof(uint64_t),
new TimestampExpectation);
}
}
void TestTimestampWritesOnRenderPass(
const std::vector<wgpu::RenderPassTimestampWrite>& timestampWrites,
const std::vector<wgpu::RenderPassTimestampWrite>& timestampWritesOnAnotherPass = {}) {
size_t queryCount = timestampWrites.size() + timestampWritesOnAnotherPass.size();
// The destination buffer offset must be a multiple of 256.
wgpu::Buffer destination =
CreateResolveBuffer(queryCount * kMinDestinationOffset + sizeof(uint64_t));
utils::BasicRenderPass renderPass = utils::CreateBasicRenderPass(device, 1, 1);
renderPass.renderPassInfo.timestampWriteCount = timestampWrites.size();
renderPass.renderPassInfo.timestampWrites = timestampWrites.data();
wgpu::CommandEncoder encoder = device.CreateCommandEncoder();
wgpu::RenderPassEncoder pass = encoder.BeginRenderPass(&renderPass.renderPassInfo);
pass.End();
// Resolve queries one by one because the query set at the beginning of pass may be
// different with the one at the end of pass.
for (size_t i = 0; i < timestampWrites.size(); i++) {
encoder.ResolveQuerySet(timestampWrites[i].querySet, timestampWrites[i].queryIndex, 1,
destination, i * kMinDestinationOffset);
}
// Begin another render pass if the timestampWritesOnAnotherPass is set.
if (!timestampWritesOnAnotherPass.empty()) {
utils::BasicRenderPass renderPass2 = utils::CreateBasicRenderPass(device, 1, 1);
renderPass2.renderPassInfo.timestampWriteCount = timestampWritesOnAnotherPass.size();
renderPass2.renderPassInfo.timestampWrites = timestampWritesOnAnotherPass.data();
wgpu::RenderPassEncoder pass2 = encoder.BeginRenderPass(&renderPass2.renderPassInfo);
pass2.End();
for (size_t i = 0; i < timestampWritesOnAnotherPass.size(); i++) {
// Resolve queries one by one because the query set at the beginning of pass may be
// different with the one at the end of pass.
encoder.ResolveQuerySet(timestampWritesOnAnotherPass[i].querySet,
timestampWritesOnAnotherPass[i].queryIndex, 1, destination,
(timestampWrites.size() + i) * kMinDestinationOffset);
}
}
wgpu::CommandBuffer commands = encoder.Finish();
queue.Submit(1, &commands);
for (size_t i = 0; i < queryCount; i++) {
EXPECT_BUFFER(destination, i * kMinDestinationOffset, sizeof(uint64_t),
new TimestampExpectation);
}
}
private:
wgpu::ComputePipeline computePipeline;
};
// Test creating query set with the type of Timestamp
@ -551,6 +669,9 @@ TEST_P(TimestampQueryTests, QuerySetCreation) {
// Test calling timestamp query from command encoder
TEST_P(TimestampQueryTests, TimestampOnCommandEncoder) {
// TODO (dawn:1250): Still not implemented on Metal backend.
DAWN_TEST_UNSUPPORTED_IF(IsMetal());
constexpr uint32_t kQueryCount = 2;
// Write timestamp with different query indexes
@ -694,6 +815,124 @@ TEST_P(TimestampQueryTests, TimestampOnComputePass) {
}
}
// Test timestampWrites setting in compute pass descriptor
TEST_P(TimestampQueryTests, TimestampWritesOnComputePass) {
// TODO(dawn:1489): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it until
// the issue is fixed.
DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsIntel());
// TODO (dawn:1250): Still not implemented on Metal backend.
DAWN_TEST_UNSUPPORTED_IF(IsMetal());
constexpr uint32_t kQueryCount = 2;
// Set timestampWrites with different query indexes and locations on same compute pass
{
wgpu::QuerySet querySet = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnComputePass(
{{querySet, 0, wgpu::ComputePassTimestampLocation::Beginning},
{querySet, 1, wgpu::ComputePassTimestampLocation::End}});
}
// Set timestampWrites with different query set on same compute pass
{
wgpu::QuerySet querySet0 = CreateQuerySetForTimestamp(1);
wgpu::QuerySet querySet1 = CreateQuerySetForTimestamp(1);
TestTimestampWritesOnComputePass(
{{querySet0, 0, wgpu::ComputePassTimestampLocation::Beginning},
{querySet1, 0, wgpu::ComputePassTimestampLocation::End}});
}
// Set timestampWrites with only one value of ComputePassTimestampLocation
{
wgpu::QuerySet querySet = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnComputePass(
{{querySet, 0, wgpu::ComputePassTimestampLocation::Beginning}});
TestTimestampWritesOnComputePass({{querySet, 1, wgpu::ComputePassTimestampLocation::End}});
}
// Set timestampWrites with same query set and query index on same compute pass
{
wgpu::QuerySet querySet = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnComputePass(
{{querySet, 0, wgpu::ComputePassTimestampLocation::Beginning},
{querySet, 0, wgpu::ComputePassTimestampLocation::End}});
}
// Set timestampWrites with same query indexes and locations on different compute pass
{
wgpu::QuerySet querySet0 = CreateQuerySetForTimestamp(kQueryCount);
wgpu::QuerySet querySet1 = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnComputePass(
{{querySet0, 0, wgpu::ComputePassTimestampLocation::Beginning},
{querySet0, 1, wgpu::ComputePassTimestampLocation::End}},
{{querySet1, 0, wgpu::ComputePassTimestampLocation::Beginning},
{querySet1, 1, wgpu::ComputePassTimestampLocation::End}});
}
}
// Test timestampWrites setting in render pass descriptor
TEST_P(TimestampQueryTests, TimestampWritesOnRenderPass) {
// TODO(dawn:1489): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it until
// the issue is fixed.
DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsIntel());
// TODO (dawn:1250): Still not implemented on Metal backend.
DAWN_TEST_UNSUPPORTED_IF(IsMetal());
constexpr uint32_t kQueryCount = 2;
// Set timestampWrites with different query indexes and locations, not need test write same
// query index due to it's not allowed on render pass.
{
wgpu::QuerySet querySet = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnRenderPass(
{{querySet, 0, wgpu::RenderPassTimestampLocation::Beginning},
{querySet, 1, wgpu::RenderPassTimestampLocation::End}});
}
// Set timestampWrites with different query set on same render pass
{
wgpu::QuerySet querySet0 = CreateQuerySetForTimestamp(1);
wgpu::QuerySet querySet1 = CreateQuerySetForTimestamp(1);
TestTimestampWritesOnRenderPass(
{{querySet0, 0, wgpu::RenderPassTimestampLocation::Beginning},
{querySet1, 0, wgpu::RenderPassTimestampLocation::End}});
}
// Set timestampWrites with only one value of RenderPassTimestampLocation
{
wgpu::QuerySet querySet = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnRenderPass(
{{querySet, 0, wgpu::RenderPassTimestampLocation::Beginning}});
TestTimestampWritesOnRenderPass({{querySet, 1, wgpu::RenderPassTimestampLocation::End}});
}
// Set timestampWrites with same query indexes and locations on different render pass
{
wgpu::QuerySet querySet0 = CreateQuerySetForTimestamp(kQueryCount);
wgpu::QuerySet querySet1 = CreateQuerySetForTimestamp(kQueryCount);
TestTimestampWritesOnRenderPass(
{{querySet0, 0, wgpu::RenderPassTimestampLocation::Beginning},
{querySet0, 1, wgpu::RenderPassTimestampLocation::End}},
{{querySet1, 0, wgpu::RenderPassTimestampLocation::Beginning},
{querySet1, 1, wgpu::RenderPassTimestampLocation::End}});
}
}
// Test resolving timestamp query from another different encoder
TEST_P(TimestampQueryTests, ResolveFromAnotherEncoder) {
constexpr uint32_t kQueryCount = 2;
@ -717,8 +956,8 @@ TEST_P(TimestampQueryTests, ResolveFromAnotherEncoder) {
// Test resolving timestamp query correctly if the queries are written sparsely
TEST_P(TimestampQueryTests, ResolveSparseQueries) {
// TODO(hao.x.li@intel.com): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it util
// TODO(dawn:1489): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it until
// the issue is fixed.
DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsIntel());
@ -768,8 +1007,8 @@ TEST_P(TimestampQueryTests, ResolveWithoutWritten) {
// Test resolving timestamp query to one slot in the buffer
TEST_P(TimestampQueryTests, ResolveToBufferWithOffset) {
// TODO(hao.x.li@intel.com): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it util
// TODO(dawn:1489): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it until
// the issue is fixed.
DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsIntel());
@ -818,8 +1057,8 @@ TEST_P(TimestampQueryTests, ResolveToBufferWithOffset) {
// Test resolving a query set twice into the same destination buffer with potentially overlapping
// ranges
TEST_P(TimestampQueryTests, ResolveTwiceToSameBuffer) {
// TODO(hao.x.li@intel.com): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it util
// TODO(dawn:1489): Fails on Intel Windows Vulkan due to a driver issue that
// vkCmdFillBuffer and vkCmdCopyQueryPoolResults are not executed in order, skip it until
// the issue is fixed.
DAWN_SUPPRESS_TEST_IF(IsWindows() && IsVulkan() && IsIntel());