griefith/test2
1
0
Fork 0
Code Issues Pull Requests Actions 1 Packages Projects Releases Wiki Activity

Metal: Improve GPU debugger support

Browse Source 
This patch improves how debug groups are displayed
within captures. Passes are now split to align with
debug groups, such that navigation of captures is
more intuitive.

To closer represent useful information, debug groups are now
deferred to align with passes, with the addition of Macros to
control capture display options.

METAL_DEBUG_CAPTURE_MAX_NESTED_GROUPS limits debug
group nesting, and METAL_DEBUG_CAPTURE_HIDE_EMPTY allows
hiding of debug groups which do not contain any commands.

Authored by Apple: Michael Parkin-White

Pull Request: https://projects.blender.org/blender/blender/pulls/108287
This commit is contained in:
Jason Fielder
2023-05-27 18:27:17 +02:00
committed by Clément Foucault
parent 52015737c9
commit 0a004fbad4
6 changed files with 155 additions and 17 deletions
Download Patch File Download Diff File Expand all files Collapse all files

2
source/blender/gpu/metal/mtl_batch.hh
View File

@@ -92,7 +92,7 @@ class MTLBatch : public Batch {
/* Returns an initialized RenderComandEncoder for drawing if all is good.
* Otherwise, nil. */
id<MTLRenderCommandEncoder> bind(uint v_count);
void unbind();
void unbind(id<MTLRenderCommandEncoder> rec);
/* Convenience getters. */
MTLIndexBuf *elem_() const

38
source/blender/gpu/metal/mtl_batch.mm
View File

@@ -480,11 +480,11 @@ id<MTLRenderCommandEncoder> MTLBatch::bind(uint v_count)
/* GPU debug markers. */
if (G.debug & G_DEBUG_GPU) {
[rec pushDebugGroup:[NSString stringWithFormat:@"batch_bind%@(shader: %s)",
[rec pushDebugGroup:[NSString stringWithFormat:@"Draw Commands%@ (GPUShader: %s)",
this->elem ? @"(indexed)" : @"",
active_shader_->get_interface()->get_name()]];
[rec insertDebugSignpost:[NSString
stringWithFormat:@"batch_bind%@(shader: %s)",
stringWithFormat:@"Draw Commands %@ (GPUShader: %s)",
this->elem ? @"(indexed)" : @"",
active_shader_->get_interface()->get_name()]];
}
@@ -557,10 +557,6 @@ id<MTLRenderCommandEncoder> MTLBatch::bind(uint v_count)
if (!ctx->ensure_render_pipeline_state(mtl_prim_type)) {
MTL_LOG_ERROR("Failed to prepare and apply render pipeline state.\n");
BLI_assert(false);
if (G.debug & G_DEBUG_GPU) {
[rec popDebugGroup];
}
return nil;
}
@@ -584,15 +580,17 @@ id<MTLRenderCommandEncoder> MTLBatch::bind(uint v_count)
rps.bind_vertex_buffer(mtl_buffer, 0, i);
}
if (G.debug & G_DEBUG_GPU) {
[rec popDebugGroup];
}
/* Return Render Command Encoder used with setup. */
return rec;
}
void MTLBatch::unbind() {}
void MTLBatch::unbind(id<MTLRenderCommandEncoder> rec)
{
/* Pop bind debug group. */
if (G.debug & G_DEBUG_GPU) {
[rec popDebugGroup];
}
}
void MTLBatch::prepare_vertex_descriptor_and_bindings(MTLVertBuf **buffers, int &num_buffers)
{
@@ -754,6 +752,8 @@ void MTLBatch::draw_advanced(int v_first, int v_count, int i_first, int i_count)
MTLContext *ctx = static_cast<MTLContext *>(unwrap(GPU_context_active_get()));
id<MTLRenderCommandEncoder> rec = this->bind(v_count);
if (rec == nil) {
/* End of draw. */
this->unbind(rec);
return;
}
@@ -885,7 +885,7 @@ void MTLBatch::draw_advanced(int v_first, int v_count, int i_first, int i_count)
}
/* End of draw. */
this->unbind();
this->unbind(rec);
}
void MTLBatch::draw_advanced_indirect(GPUStorageBuf *indirect_buf, intptr_t offset)
@@ -895,6 +895,9 @@ void MTLBatch::draw_advanced_indirect(GPUStorageBuf *indirect_buf, intptr_t offs
id<MTLRenderCommandEncoder> rec = this->bind(0);
if (rec == nil) {
printf("Failed to open Render Command encoder for DRAW INDIRECT\n");
/* End of draw. */
this->unbind(rec);
return;
}
@@ -902,6 +905,9 @@ void MTLBatch::draw_advanced_indirect(GPUStorageBuf *indirect_buf, intptr_t offs
* NOTE: Add support? */
if (active_shader_->get_uses_ssbo_vertex_fetch()) {
printf("Draw indirect for SSBO vertex fetch disabled\n");
/* End of draw. */
this->unbind(rec);
return;
}
@@ -911,6 +917,9 @@ void MTLBatch::draw_advanced_indirect(GPUStorageBuf *indirect_buf, intptr_t offs
if (mtl_needs_topology_emulation(this->prim_type)) {
BLI_assert_msg(false, "Metal Topology emulation unsupported for draw indirect.\n");
/* End of draw. */
this->unbind(rec);
return;
}
@@ -920,6 +929,9 @@ void MTLBatch::draw_advanced_indirect(GPUStorageBuf *indirect_buf, intptr_t offs
BLI_assert(mtl_indirect_buf != nil);
if (mtl_indirect_buf == nil) {
MTL_LOG_WARNING("Metal Indirect Draw Storage Buffer is nil.\n");
/* End of draw. */
this->unbind(rec);
return;
}
@@ -963,7 +975,7 @@ void MTLBatch::draw_advanced_indirect(GPUStorageBuf *indirect_buf, intptr_t offs
}
/* End of draw. */
this->unbind();
this->unbind(rec);
}
/** \} */

108
source/blender/gpu/metal/mtl_command_buffer.mm
View File

@@ -75,6 +75,10 @@ id<MTLCommandBuffer> MTLCommandBufferManager::ensure_begin()
/* Reset Command buffer heuristics. */
this->reset_counters();
/* Clear debug stacks. */
debug_group_stack.clear();
debug_group_pushed_stack.clear();
}
BLI_assert(active_command_buffer_ != nil);
return active_command_buffer_;
@@ -322,12 +326,27 @@ id<MTLRenderCommandEncoder> MTLCommandBufferManager::ensure_begin_render_command
/* Ensure we have already cleaned up our previous render command encoder. */
BLI_assert(active_render_command_encoder_ == nil);
/* Unroll pending debug groups. */
if (G.debug & G_DEBUG_GPU) {
unfold_pending_debug_groups();
}
/* Create new RenderCommandEncoder based on descriptor (and begin encoding). */
active_render_command_encoder_ = [cmd_buf
renderCommandEncoderWithDescriptor:active_pass_descriptor_];
[active_render_command_encoder_ retain];
active_command_encoder_type_ = MTL_RENDER_COMMAND_ENCODER;
/* Add debug label. */
if (G.debug & G_DEBUG_GPU) {
std::string debug_name = "RenderCmdEncoder: Unnamed";
if (!debug_group_pushed_stack.empty()) {
debug_name = "RenderCmdEncoder: " + debug_group_pushed_stack.back();
}
debug_name += " (FrameBuffer: " + std::string(active_frame_buffer_->name_get()) + ")";
active_render_command_encoder_.label = [NSString stringWithUTF8String:debug_name.c_str()];
}
/* Update command buffer encoder heuristics. */
this->register_encoder_counters();
@@ -367,11 +386,25 @@ id<MTLBlitCommandEncoder> MTLCommandBufferManager::ensure_begin_blit_encoder()
/* Begin new Blit Encoder. */
if (active_blit_command_encoder_ == nil) {
/* Unroll pending debug groups. */
if (G.debug & G_DEBUG_GPU) {
unfold_pending_debug_groups();
}
active_blit_command_encoder_ = [cmd_buf blitCommandEncoder];
BLI_assert(active_blit_command_encoder_ != nil);
[active_blit_command_encoder_ retain];
active_command_encoder_type_ = MTL_BLIT_COMMAND_ENCODER;
/* Add debug label. */
if (G.debug & G_DEBUG_GPU) {
std::string debug_name = "BlitCmdEncoder: Unnamed";
if (!debug_group_pushed_stack.empty()) {
debug_name = "BlitCmdEncoder: " + debug_group_pushed_stack.back();
}
active_blit_command_encoder_.label = [NSString stringWithUTF8String:debug_name.c_str()];
}
/* Update command buffer encoder heuristics. */
this->register_encoder_counters();
}
@@ -392,11 +425,25 @@ id<MTLComputeCommandEncoder> MTLCommandBufferManager::ensure_begin_compute_encod
/* Begin new Compute Encoder. */
if (active_compute_command_encoder_ == nil) {
/* Unroll pending debug groups. */
if (G.debug & G_DEBUG_GPU) {
unfold_pending_debug_groups();
}
active_compute_command_encoder_ = [cmd_buf computeCommandEncoder];
BLI_assert(active_compute_command_encoder_ != nil);
[active_compute_command_encoder_ retain];
active_command_encoder_type_ = MTL_COMPUTE_COMMAND_ENCODER;
/* Add debug label. */
if (G.debug & G_DEBUG_GPU) {
std::string debug_name = "ComputeCmdEncoder: Unnamed";
if (!debug_group_pushed_stack.empty()) {
debug_name = "ComputeCmdEncoder: " + debug_group_pushed_stack.back();
}
active_compute_command_encoder_.label = [NSString stringWithUTF8String:debug_name.c_str()];
}
/* Update command buffer encoder heuristics. */
this->register_encoder_counters();
@@ -461,17 +508,74 @@ bool MTLCommandBufferManager::do_break_submission()
/* Debug. */
void MTLCommandBufferManager::push_debug_group(const char *name, int index)
{
/* Only perform this operation if capturing. */
MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager];
if (![capture_manager isCapturing]) {
return;
}
id<MTLCommandBuffer> cmd = this->ensure_begin();
if (cmd != nil) {
[cmd pushDebugGroup:[NSString stringWithFormat:@"%s_%d", name, index]];
if (active_command_encoder_type_ != MTL_NO_COMMAND_ENCODER) {
end_active_command_encoder();
}
debug_group_stack.push_back(std::string(name));
}
}
void MTLCommandBufferManager::pop_debug_group()
{
/* Only perform this operation if capturing. */
MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager];
if (![capture_manager isCapturing]) {
return;
}
id<MTLCommandBuffer> cmd = this->ensure_begin();
if (cmd != nil) {
[cmd popDebugGroup];
if (active_command_encoder_type_ != MTL_NO_COMMAND_ENCODER) {
end_active_command_encoder();
}
#if METAL_DEBUG_CAPTURE_HIDE_EMPTY == 0
/* Unfold pending groups to display empty groups. */
unfold_pending_debug_groups();
#endif
/* If we have pending debug groups, first pop the last pending one. */
if (debug_group_stack.size() > 0) {
debug_group_stack.pop_back();
}
else {
/* Otherwise, close last active pushed group. */
if (debug_group_pushed_stack.size() > 0) {
debug_group_pushed_stack.pop_back();
if (debug_group_pushed_stack.size() < uint(METAL_DEBUG_CAPTURE_MAX_NESTED_GROUPS)) {
[cmd popDebugGroup];
}
}
}
}
}
void MTLCommandBufferManager::unfold_pending_debug_groups()
{
/* Only perform this operation if capturing. */
MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager];
if (![capture_manager isCapturing]) {
return;
}
if (active_command_buffer_ != nil) {
for (const std::string &name : debug_group_stack) {
if (debug_group_pushed_stack.size() < uint(METAL_DEBUG_CAPTURE_MAX_NESTED_GROUPS)) {
[active_command_buffer_ pushDebugGroup:[NSString stringWithFormat:@"%s", name.c_str()]];
}
debug_group_pushed_stack.push_back(name);
}
debug_group_stack.clear();
}
}

10
source/blender/gpu/metal/mtl_context.hh
View File

@@ -575,6 +575,13 @@ class MTLCommandBufferManager {
int vertex_submitted_count_ = 0;
bool empty_ = true;
/** Debug groups. */
/* Stack tracking all calls to push_debug_group. */
std::vector<std::string> debug_group_stack;
/* Stack tracking calls resulting in active API calls to pushDebugGroup on the current command
* buffer. */
std::vector<std::string> debug_group_pushed_stack;
public:
MTLCommandBufferManager(MTLContext &context)
: context_(context), render_pass_state_(context, *this), compute_state_(context, *this){};
@@ -639,6 +646,9 @@ class MTLCommandBufferManager {
id<MTLCommandBuffer> ensure_begin();
void register_encoder_counters();
/* Debug group management. */
void unfold_pending_debug_groups();
};
/** MTLContext -- Core render loop and state management. **/

9
source/blender/gpu/metal/mtl_debug.hh
View File

@@ -9,6 +9,15 @@
#include "BKE_global.h"
#include "CLG_log.h"
/** Options for organising Metal GPU debug captures. */
/* Maximum nested debug group depth. Groups beyond this will still have the pass name pulled into
* the RenderCommandEncoder, but will not display in the trace.
* Use -1 for unlimited. */
#define METAL_DEBUG_CAPTURE_MAX_NESTED_GROUPS -1
/* Whether empty debug groups should be hidden. */
#define METAL_DEBUG_CAPTURE_HIDE_EMPTY 0
namespace blender {
namespace gpu {
namespace debug {

5
source/blender/gpu/metal/mtl_drawlist.mm
View File

@@ -189,6 +189,9 @@ void MTLDrawList::submit()
id<MTLRenderCommandEncoder> rec = batch_->bind(0);
if (rec == nil) {
BLI_assert_msg(false, "A RenderCommandEncoder should always be available!\n");
/* Unbind batch. */
batch_->unbind(rec);
return;
}
@@ -272,7 +275,7 @@ void MTLDrawList::submit()
}
/* Unbind batch. */
batch_->unbind();
batch_->unbind(rec);
/* Reset command offsets. */
command_len_ = 0;