diff options
28 files changed, 3868 insertions, 731 deletions
diff --git a/source/blender/gpu/CMakeLists.txt b/source/blender/gpu/CMakeLists.txt index cadc2c4445b..d8ed74390f4 100644 --- a/source/blender/gpu/CMakeLists.txt +++ b/source/blender/gpu/CMakeLists.txt @@ -188,7 +188,9 @@ set(OPENGL_SRC set(METAL_SRC metal/mtl_backend.mm metal/mtl_context.mm + metal/mtl_command_buffer.mm metal/mtl_debug.mm + metal/mtl_framebuffer.mm metal/mtl_state.mm metal/mtl_texture.mm metal/mtl_texture_util.mm @@ -198,6 +200,7 @@ set(METAL_SRC metal/mtl_common.hh metal/mtl_context.hh metal/mtl_debug.hh + metal/mtl_framebuffer.hh metal/mtl_state.hh metal/mtl_texture.hh ) diff --git a/source/blender/gpu/GPU_common_types.h b/source/blender/gpu/GPU_common_types.h index 8c91d60812f..5913caf72e3 100644 --- a/source/blender/gpu/GPU_common_types.h +++ b/source/blender/gpu/GPU_common_types.h @@ -8,6 +8,14 @@ extern "C" { #endif +typedef enum eGPULoadOp { + GPU_LOADACTION_CLEAR = 0, + GPU_LOADACTION_LOAD, + GPU_LOADACTION_DONT_CARE +} eGPULoadOp; + +typedef enum eGPUStoreOp { GPU_STOREACTION_STORE = 0, GPU_STOREACTION_DONT_CARE } eGPUStoreOp; + typedef enum eGPUFrontFace { GPU_CLOCKWISE, GPU_COUNTERCLOCKWISE, diff --git a/source/blender/gpu/GPU_context.h b/source/blender/gpu/GPU_context.h index f3b7f8c29bf..1fcd94c48fc 100644 --- a/source/blender/gpu/GPU_context.h +++ b/source/blender/gpu/GPU_context.h @@ -38,6 +38,13 @@ void GPU_context_discard(GPUContext *); void GPU_context_active_set(GPUContext *); GPUContext *GPU_context_active_get(void); +/* Begin and end frame are used to mark the singular boundary representing the lifetime of a whole + * frame. This also acts as a divisor for ensuring workload submission and flushing, especially for + * background rendering when there is no call to present. + * This is required by explicit-API's where there is no implicit workload flushing. */ +void GPU_context_begin_frame(GPUContext *ctx); +void GPU_context_end_frame(GPUContext *ctx); + /* Legacy GPU (Intel HD4000 series) do not support sharing GPU objects between GPU * contexts. EEVEE/Workbench can create different contexts for image/preview rendering, baking or * compiling. When a legacy GPU is detected (`GPU_use_main_context_workaround()`) any worker diff --git a/source/blender/gpu/GPU_framebuffer.h b/source/blender/gpu/GPU_framebuffer.h index 4436f7a5a7b..6eb51c200f1 100644 --- a/source/blender/gpu/GPU_framebuffer.h +++ b/source/blender/gpu/GPU_framebuffer.h @@ -14,6 +14,7 @@ #pragma once +#include "GPU_common_types.h" #include "GPU_texture.h" typedef enum eGPUFrameBufferBits { @@ -52,6 +53,44 @@ void GPU_framebuffer_bind(GPUFrameBuffer *fb); void GPU_framebuffer_bind_no_srgb(GPUFrameBuffer *fb); void GPU_framebuffer_restore(void); +/* Advanced binding control. */ +typedef struct GPULoadStore { + eGPULoadOp load_action; + eGPUStoreOp store_action; +} GPULoadStore; +#define NULL_LOAD_STORE \ + { \ + GPU_LOADACTION_DONT_CARE, GPU_STOREACTION_DONT_CARE \ + } + +/* Load store config array (load_store_actions) matches attachment structure of + * GPU_framebuffer_config_array. This allows us to explicitly specify whether attachment data needs + * to be loaded and stored on a per-attachment basis. This enables a number of bandwidth + * optimisations: + * - No need to load contents if subsequent work is over-writing every pixel. + * - No need to store attachments whose contents are not used beyond this pass e.g. depth buffer. + * - State can be customised at bind-time rather than applying to the framebuffer object as a + * whole. + * + * Example: + * \code{.c} + * GPU_framebuffer_bind_loadstore(&fb, { + * {GPU_LOADACTION_LOAD, GPU_STOREACTION_DONT_CARE} // must be depth buffer + * {GPU_LOADACTION_LOAD, GPU_STOREACTION_STORE}, // Colour attachment 0 + * {GPU_LOADACTION_DONT_CARE, GPU_STOREACTION_STORE}, // Colour attachment 1 + * {GPU_LOADACTION_DONT_CARE, GPU_STOREACTION_STORE} // Colour attachment 2 + * }) + * \encode + */ +void GPU_framebuffer_bind_loadstore(GPUFrameBuffer *fb, + const GPULoadStore *load_store_actions, + uint actions_len); +#define GPU_framebuffer_bind_ex(_fb, ...) \ + { \ + GPULoadStore actions[] = __VA_ARGS__; \ + GPU_framebuffer_bind_loadstore(_fb, actions, (sizeof(actions) / sizeof(GPULoadStore))); \ + } + bool GPU_framebuffer_bound(GPUFrameBuffer *fb); bool GPU_framebuffer_check_valid(GPUFrameBuffer *fb, char err_out[256]); diff --git a/source/blender/gpu/intern/gpu_context.cc b/source/blender/gpu/intern/gpu_context.cc index 9fb5826506a..4a0a9ecc7f6 100644 --- a/source/blender/gpu/intern/gpu_context.cc +++ b/source/blender/gpu/intern/gpu_context.cc @@ -123,6 +123,22 @@ GPUContext *GPU_context_active_get() return wrap(Context::get()); } +void GPU_context_begin_frame(GPUContext *ctx) +{ + blender::gpu::Context *_ctx = unwrap(ctx); + if (_ctx) { + _ctx->begin_frame(); + } +} + +void GPU_context_end_frame(GPUContext *ctx) +{ + blender::gpu::Context *_ctx = unwrap(ctx); + if (_ctx) { + _ctx->end_frame(); + } +} + /* -------------------------------------------------------------------- */ /** \name Main context global mutex * diff --git a/source/blender/gpu/intern/gpu_context_private.hh b/source/blender/gpu/intern/gpu_context_private.hh index af9791fde88..9cdf0075632 100644 --- a/source/blender/gpu/intern/gpu_context_private.hh +++ b/source/blender/gpu/intern/gpu_context_private.hh @@ -63,6 +63,8 @@ class Context { virtual void activate() = 0; virtual void deactivate() = 0; + virtual void begin_frame() = 0; + virtual void end_frame() = 0; /* Will push all pending commands to the GPU. */ virtual void flush() = 0; diff --git a/source/blender/gpu/intern/gpu_framebuffer.cc b/source/blender/gpu/intern/gpu_framebuffer.cc index 08d761106e5..f12d8fd7e55 100644 --- a/source/blender/gpu/intern/gpu_framebuffer.cc +++ b/source/blender/gpu/intern/gpu_framebuffer.cc @@ -124,6 +124,43 @@ void FrameBuffer::attachment_remove(GPUAttachmentType type) dirty_attachments_ = true; } +void FrameBuffer::load_store_config_array(const GPULoadStore *load_store_actions, uint actions_len) +{ + /* Follows attachment structure of GPU_framebuffer_config_array/GPU_framebuffer_ensure_config */ + const GPULoadStore &depth_action = load_store_actions[0]; + Span<GPULoadStore> color_attachments(load_store_actions + 1, actions_len - 1); + + if (this->attachments_[GPU_FB_DEPTH_STENCIL_ATTACHMENT].tex) { + this->attachment_set_loadstore_op( + GPU_FB_DEPTH_STENCIL_ATTACHMENT, depth_action.load_action, depth_action.store_action); + } + if (this->attachments_[GPU_FB_DEPTH_ATTACHMENT].tex) { + this->attachment_set_loadstore_op( + GPU_FB_DEPTH_ATTACHMENT, depth_action.load_action, depth_action.store_action); + } + + GPUAttachmentType type = GPU_FB_COLOR_ATTACHMENT0; + for (const GPULoadStore &actions : color_attachments) { + if (this->attachments_[type].tex) { + this->attachment_set_loadstore_op(type, actions.load_action, actions.store_action); + } + ++type; + } +} + +unsigned int FrameBuffer::get_bits_per_pixel(void) +{ + unsigned int total_bits = 0; + for (GPUAttachment &attachment : attachments_) { + Texture *tex = reinterpret_cast<Texture *>(attachment.tex); + if (tex != nullptr) { + int bits = to_bytesize(tex->format_get()) * to_component_len(tex->format_get()); + total_bits += bits; + } + } + return total_bits; +} + void FrameBuffer::recursive_downsample(int max_lvl, void (*callback)(void *userData, int level), void *userData) @@ -149,10 +186,21 @@ void FrameBuffer::recursive_downsample(int max_lvl, attachment.mip = mip_lvl; } } + /* Update the internal attachments and viewport size. */ dirty_attachments_ = true; this->bind(true); + /* Optimise load-store state. */ + GPUAttachmentType type = GPU_FB_DEPTH_ATTACHMENT; + for (GPUAttachment &attachment : attachments_) { + Texture *tex = reinterpret_cast<Texture *>(attachment.tex); + if (tex != nullptr) { + this->attachment_set_loadstore_op(type, GPU_LOADACTION_DONT_CARE, GPU_STOREACTION_STORE); + } + ++type; + } + callback(userData, mip_lvl); } @@ -198,6 +246,18 @@ void GPU_framebuffer_bind(GPUFrameBuffer *gpu_fb) unwrap(gpu_fb)->bind(enable_srgb); } +void GPU_framebuffer_bind_loadstore(GPUFrameBuffer *gpu_fb, + const GPULoadStore *load_store_actions, + uint actions_len) +{ + /* Bind */ + GPU_framebuffer_bind(gpu_fb); + + /* Update load store */ + FrameBuffer *fb = unwrap(gpu_fb); + fb->load_store_config_array(load_store_actions, actions_len); +} + void GPU_framebuffer_bind_no_srgb(GPUFrameBuffer *gpu_fb) { const bool enable_srgb = false; diff --git a/source/blender/gpu/intern/gpu_framebuffer_private.hh b/source/blender/gpu/intern/gpu_framebuffer_private.hh index d218662d17f..8cecc6b8b15 100644 --- a/source/blender/gpu/intern/gpu_framebuffer_private.hh +++ b/source/blender/gpu/intern/gpu_framebuffer_private.hh @@ -114,6 +114,10 @@ class FrameBuffer { eGPUDataFormat data_format, const void *clear_value) = 0; + virtual void attachment_set_loadstore_op(GPUAttachmentType type, + eGPULoadOp load_action, + eGPUStoreOp store_action) = 0; + virtual void read(eGPUFrameBufferBits planes, eGPUDataFormat format, const int area[4], @@ -128,12 +132,15 @@ class FrameBuffer { int dst_offset_x, int dst_offset_y) = 0; + void load_store_config_array(const GPULoadStore *load_store_actions, uint actions_len); + void attachment_set(GPUAttachmentType type, const GPUAttachment &new_attachment); void attachment_remove(GPUAttachmentType type); void recursive_downsample(int max_lvl, void (*callback)(void *userData, int level), void *userData); + uint get_bits_per_pixel(); inline void size_set(int width, int height) { diff --git a/source/blender/gpu/metal/mtl_backend.hh b/source/blender/gpu/metal/mtl_backend.hh index 9044d8517ab..7228a5f7596 100644 --- a/source/blender/gpu/metal/mtl_backend.hh +++ b/source/blender/gpu/metal/mtl_backend.hh @@ -35,19 +35,19 @@ class MTLBackend : public GPUBackend { return MTLBackend::capabilities; } - inline ~MTLBackend() + ~MTLBackend() { MTLBackend::platform_exit(); } static bool metal_is_supported(); - inline static MTLBackend *get() + static MTLBackend *get() { return static_cast<MTLBackend *>(GPUBackend::get()); } void samplers_update() override; - inline void compute_dispatch(int groups_x_len, int groups_y_len, int groups_z_len) override + void compute_dispatch(int groups_x_len, int groups_y_len, int groups_z_len) override { /* Placeholder */ } diff --git a/source/blender/gpu/metal/mtl_backend.mm b/source/blender/gpu/metal/mtl_backend.mm index 1064091a036..81f8f279759 100644 --- a/source/blender/gpu/metal/mtl_backend.mm +++ b/source/blender/gpu/metal/mtl_backend.mm @@ -9,6 +9,7 @@ #include "gpu_backend.hh" #include "mtl_backend.hh" #include "mtl_context.hh" +#include "mtl_framebuffer.hh" #include "gpu_capabilities_private.hh" #include "gpu_platform_private.hh" @@ -50,8 +51,9 @@ DrawList *MTLBackend::drawlist_alloc(int list_length) FrameBuffer *MTLBackend::framebuffer_alloc(const char *name) { - /* TODO(Metal): Implement MTLFrameBuffer. */ - return nullptr; + MTLContext *mtl_context = static_cast<MTLContext *>( + reinterpret_cast<Context *>(GPU_context_active_get())); + return new MTLFrameBuffer(mtl_context, name); }; IndexBuf *MTLBackend::indexbuf_alloc() @@ -380,11 +382,10 @@ void MTLBackend::capabilities_init(MTLContext *ctx) /* In Metal, total_thread_count is 512 or 1024, such that * threadgroup `width*height*depth <= total_thread_count` */ - unsigned int max_threads_per_threadgroup_per_dim = - ([device supportsFamily:MTLGPUFamilyApple4] || - MTLBackend::capabilities.supports_family_mac1) ? - 1024 : - 512; + uint max_threads_per_threadgroup_per_dim = ([device supportsFamily:MTLGPUFamilyApple4] || + MTLBackend::capabilities.supports_family_mac1) ? + 1024 : + 512; GCaps.max_work_group_size[0] = max_threads_per_threadgroup_per_dim; GCaps.max_work_group_size[1] = max_threads_per_threadgroup_per_dim; GCaps.max_work_group_size[2] = max_threads_per_threadgroup_per_dim; diff --git a/source/blender/gpu/metal/mtl_capabilities.hh b/source/blender/gpu/metal/mtl_capabilities.hh index 3afa6e31ccb..d56f796e60f 100644 --- a/source/blender/gpu/metal/mtl_capabilities.hh +++ b/source/blender/gpu/metal/mtl_capabilities.hh @@ -19,7 +19,7 @@ namespace gpu { #define MTL_MAX_UNIFORMS_PER_BLOCK 64 /* Context-specific limits -- populated in 'MTLBackend::platform_init' */ -typedef struct MTLCapabilities { +struct MTLCapabilities { /* Variable Limits & feature sets. */ int max_color_render_targets = 4; /* Minimum = 4 */ @@ -40,8 +40,7 @@ typedef struct MTLCapabilities { bool supports_family_mac2 = false; bool supports_family_mac_catalyst1 = false; bool supports_family_mac_catalyst2 = false; - -} MTLCapabilities; +}; } // namespace gpu } // namespace blender diff --git a/source/blender/gpu/metal/mtl_command_buffer.mm b/source/blender/gpu/metal/mtl_command_buffer.mm new file mode 100644 index 00000000000..392e07b7ee8 --- /dev/null +++ b/source/blender/gpu/metal/mtl_command_buffer.mm @@ -0,0 +1,635 @@ + +#include "DNA_userdef_types.h" + +#include "mtl_backend.hh" +#include "mtl_common.hh" +#include "mtl_context.hh" +#include "mtl_debug.hh" +#include "mtl_framebuffer.hh" + +#include <fstream> + +using namespace blender; +using namespace blender::gpu; + +namespace blender::gpu { + +/* Global sync event used across MTLContext's. + * This resolves flickering artifacts from command buffer + * dependencies not being honoured for work submitted between + * different GPUContext's. */ +id<MTLEvent> MTLCommandBufferManager::sync_event = nil; +unsigned long long MTLCommandBufferManager::event_signal_val = 0; + +/* Counter for active comand buffers. */ +int MTLCommandBufferManager::num_active_cmd_bufs = 0; + +/* -------------------------------------------------------------------- */ +/** \name MTLCommandBuffer initialisation and render coordination. + * \{ */ + +void MTLCommandBufferManager::prepare(MTLContext *ctx, bool supports_render) +{ + context_ = ctx; + render_pass_state_.prepare(this, ctx); +} + +void MTLCommandBufferManager::register_encoder_counters() +{ + encoder_count_++; + empty_ = false; +} + +id<MTLCommandBuffer> MTLCommandBufferManager::ensure_begin() +{ + if (active_command_buffer_ == nil) { + + /* Verify number of active command buffers is below limit. + * Exceeding this limit will mean we either have a leak/GPU hang + * or we should increase the command buffer limit during MTLQueue creation */ + BLI_assert(MTLCommandBufferManager::num_active_cmd_bufs < MTL_MAX_COMMAND_BUFFERS); + + if (G.debug & G_DEBUG_GPU) { + /* Debug: Enable Advanced Errors for GPU work execution. */ + MTLCommandBufferDescriptor *desc = [[MTLCommandBufferDescriptor alloc] init]; + desc.errorOptions = MTLCommandBufferErrorOptionEncoderExecutionStatus; + desc.retainedReferences = YES; + active_command_buffer_ = [context_->queue commandBufferWithDescriptor:desc]; + } + else { + active_command_buffer_ = [context_->queue commandBuffer]; + } + [active_command_buffer_ retain]; + MTLCommandBufferManager::num_active_cmd_bufs++; + + /* Ensure command buffers execute in submission order across multiple MTLContext's. */ + if (this->sync_event != nil) { + [active_command_buffer_ encodeWaitForEvent:this->sync_event value:this->event_signal_val]; + } + + /* Reset Command buffer heuristics. */ + this->reset_counters(); + } + BLI_assert(active_command_buffer_ != nil); + return active_command_buffer_; +} + +/* If wait is true, CPU will stall until GPU work has completed. */ +bool MTLCommandBufferManager::submit(bool wait) +{ + /* Skip submission if command buffer is empty. */ + if (empty_ || active_command_buffer_ == nil) { + return false; + } + + /* Ensure current encoders are finished. */ + this->end_active_command_encoder(); + BLI_assert(active_command_encoder_type_ == MTL_NO_COMMAND_ENCODER); + + /*** Submit Command Buffer. ***/ + /* Strict ordering ensures command buffers are guaranteed to execute after a previous + * one has completed. Resolves flickering when command buffers are submitted from + * different MTLContext's. */ + if (MTLCommandBufferManager::sync_event == nil) { + MTLCommandBufferManager::sync_event = [context_->device newEvent]; + BLI_assert(MTLCommandBufferManager::sync_event); + [MTLCommandBufferManager::sync_event retain]; + } + BLI_assert(MTLCommandBufferManager::sync_event != nil); + MTLCommandBufferManager::event_signal_val++; + + [active_command_buffer_ encodeSignalEvent:MTLCommandBufferManager::sync_event + value:MTLCommandBufferManager::event_signal_val]; + + /* Command buffer lifetime tracking. */ + /* TODO(Metal): This routine will later be used to track released memory allocations within the + * lifetime of a command buffer such that memory is only released once no longer in use. */ + id<MTLCommandBuffer> cmd_buffer_ref = [active_command_buffer_ retain]; + [cmd_buffer_ref addCompletedHandler:^(id<MTLCommandBuffer> cb) { + /* Release command buffer after completion callback handled. */ + [cmd_buffer_ref release]; + + /* Decrement active cmd buffer count. */ + MTLCommandBufferManager::num_active_cmd_bufs--; + }]; + + /* Submit command buffer to GPU. */ + [active_command_buffer_ commit]; + + if (wait || (G.debug & G_DEBUG_GPU)) { + /* Wait until current GPU work has finished executing. */ + [active_command_buffer_ waitUntilCompleted]; + + /* Command buffer execution debugging can return an error message if + * execution has failed or encoutered GPU-side errors. */ + if (G.debug & G_DEBUG_GPU) { + + NSError *error = [active_command_buffer_ error]; + if (error != nil) { + NSLog(@"%@", error); + BLI_assert(false); + + @autoreleasepool { + const char *stringAsChar = [[NSString stringWithFormat:@"%@", error] UTF8String]; + + std::ofstream outfile; + outfile.open("command_buffer_error.txt", std::fstream::out | std::fstream::app); + outfile << stringAsChar; + outfile.close(); + } + } + } + } + + /* Release previous frames command buffer and reset active cmd buffer. */ + if (last_submitted_command_buffer_ != nil) { + + BLI_assert(MTLBackend::get()->is_inside_render_boundary()); + [last_submitted_command_buffer_ autorelease]; + last_submitted_command_buffer_ = nil; + } + last_submitted_command_buffer_ = active_command_buffer_; + active_command_buffer_ = nil; + + return true; +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \name Render Command Encoder Utility and management functions. + * \{ */ + +/* Fetch/query current encoder. */ +bool MTLCommandBufferManager::is_inside_render_pass() +{ + return (active_command_encoder_type_ == MTL_RENDER_COMMAND_ENCODER); +} + +bool MTLCommandBufferManager::is_inside_blit() +{ + return (active_command_encoder_type_ == MTL_BLIT_COMMAND_ENCODER); +} + +bool MTLCommandBufferManager::is_inside_compute() +{ + return (active_command_encoder_type_ == MTL_COMPUTE_COMMAND_ENCODER); +} + +id<MTLRenderCommandEncoder> MTLCommandBufferManager::get_active_render_command_encoder() +{ + /* Calling code should check if inside render pass. Otherwise nil. */ + return active_render_command_encoder_; +} + +id<MTLBlitCommandEncoder> MTLCommandBufferManager::get_active_blit_command_encoder() +{ + /* Calling code should check if inside render pass. Otherwise nil. */ + return active_blit_command_encoder_; +} + +id<MTLComputeCommandEncoder> MTLCommandBufferManager::get_active_compute_command_encoder() +{ + /* Calling code should check if inside render pass. Otherwise nil. */ + return active_compute_command_encoder_; +} + +MTLFrameBuffer *MTLCommandBufferManager::get_active_framebuffer() +{ + /* If outside of RenderPass, nullptr will be returned. */ + if (this->is_inside_render_pass()) { + return active_frame_buffer_; + } + return nullptr; +} + +/* Encoder and Pass management. */ +/* End currently active MTLCommandEncoder. */ +bool MTLCommandBufferManager::end_active_command_encoder() +{ + + /* End active encoder if one is active. */ + if (active_command_encoder_type_ != MTL_NO_COMMAND_ENCODER) { + + switch (active_command_encoder_type_) { + case MTL_RENDER_COMMAND_ENCODER: { + /* Verify a RenderCommandEncoder is active and end. */ + BLI_assert(active_render_command_encoder_ != nil); + + /* Complete Encoding. */ + [active_render_command_encoder_ endEncoding]; + [active_render_command_encoder_ release]; + active_render_command_encoder_ = nil; + active_command_encoder_type_ = MTL_NO_COMMAND_ENCODER; + + /* Reset associated framebuffer flag. */ + active_frame_buffer_ = nullptr; + active_pass_descriptor_ = nullptr; + return true; + } + + case MTL_BLIT_COMMAND_ENCODER: { + /* Verify a RenderCommandEncoder is active and end. */ + BLI_assert(active_blit_command_encoder_ != nil); + [active_blit_command_encoder_ endEncoding]; + [active_blit_command_encoder_ release]; + active_blit_command_encoder_ = nil; + active_command_encoder_type_ = MTL_NO_COMMAND_ENCODER; + return true; + } + + case MTL_COMPUTE_COMMAND_ENCODER: { + /* Verify a RenderCommandEncoder is active and end. */ + BLI_assert(active_compute_command_encoder_ != nil); + [active_compute_command_encoder_ endEncoding]; + [active_compute_command_encoder_ release]; + active_compute_command_encoder_ = nil; + active_command_encoder_type_ = MTL_NO_COMMAND_ENCODER; + return true; + } + + default: { + BLI_assert(false && "Invalid command encoder type"); + return false; + } + }; + } + else { + /* MTL_NO_COMMAND_ENCODER. */ + BLI_assert(active_render_command_encoder_ == nil); + BLI_assert(active_blit_command_encoder_ == nil); + BLI_assert(active_compute_command_encoder_ == nil); + return false; + } +} + +id<MTLRenderCommandEncoder> MTLCommandBufferManager::ensure_begin_render_command_encoder( + MTLFrameBuffer *ctx_framebuffer, bool force_begin, bool *new_pass) +{ + /* Ensure valid framebuffer. */ + BLI_assert(ctx_framebuffer != nullptr); + + /* Ensure active command buffer. */ + id<MTLCommandBuffer> cmd_buf = this->ensure_begin(); + BLI_assert(cmd_buf); + + /* Begin new command encoder if the currently active one is + * incompatible or requires updating. */ + if (active_command_encoder_type_ != MTL_RENDER_COMMAND_ENCODER || + active_frame_buffer_ != ctx_framebuffer || force_begin) { + this->end_active_command_encoder(); + + /* Determine if this is a re-bind of the same framebuffer. */ + bool is_rebind = (active_frame_buffer_ == ctx_framebuffer); + + /* Generate RenderPassDescriptor from bound framebuffer. */ + BLI_assert(ctx_framebuffer); + active_frame_buffer_ = ctx_framebuffer; + active_pass_descriptor_ = active_frame_buffer_->bake_render_pass_descriptor( + is_rebind && (!active_frame_buffer_->get_pending_clear())); + + /* Ensure we have already cleaned up our previous render command encoder. */ + BLI_assert(active_render_command_encoder_ == nil); + + /* 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; + + /* Update command buffer encoder heuristics. */ + this->register_encoder_counters(); + + /* Apply initial state. */ + /* Update Viewport and Scissor State */ + active_frame_buffer_->apply_state(); + + /* FLAG FRAMEBUFFER AS CLEARED -- A clear only lasts as long as one has been specified. + * After this, resets to Load attachments to parallel GL behaviour. */ + active_frame_buffer_->mark_cleared(); + + /* Reset RenderPassState to ensure resource bindings are re-applied. */ + render_pass_state_.reset_state(); + + /* Return true as new pass started. */ + *new_pass = true; + } + else { + /* No new pass. */ + *new_pass = false; + } + + BLI_assert(active_render_command_encoder_ != nil); + return active_render_command_encoder_; +} + +id<MTLBlitCommandEncoder> MTLCommandBufferManager::ensure_begin_blit_encoder() +{ + /* Ensure active command buffer. */ + id<MTLCommandBuffer> cmd_buf = this->ensure_begin(); + BLI_assert(cmd_buf); + + /* Ensure no existing command encoder of a different type is active. */ + if (active_command_encoder_type_ != MTL_BLIT_COMMAND_ENCODER) { + this->end_active_command_encoder(); + } + + /* Begin new Blit Encoder. */ + if (active_blit_command_encoder_ == nil) { + 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; + + /* Update command buffer encoder heuristics. */ + this->register_encoder_counters(); + } + BLI_assert(active_blit_command_encoder_ != nil); + return active_blit_command_encoder_; +} + +id<MTLComputeCommandEncoder> MTLCommandBufferManager::ensure_begin_compute_encoder() +{ + /* Ensure active command buffer. */ + id<MTLCommandBuffer> cmd_buf = this->ensure_begin(); + BLI_assert(cmd_buf); + + /* Ensure no existing command encoder of a different type is active. */ + if (active_command_encoder_type_ != MTL_COMPUTE_COMMAND_ENCODER) { + this->end_active_command_encoder(); + } + + /* Begin new Compute Encoder. */ + if (active_compute_command_encoder_ == nil) { + 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; + + /* Update command buffer encoder heuristics. */ + this->register_encoder_counters(); + } + BLI_assert(active_compute_command_encoder_ != nil); + return active_compute_command_encoder_; +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \name Command buffer heuristics. + * \{ */ + +/* Rendering Heuristics. */ +void MTLCommandBufferManager::register_draw_counters(int vertex_submission) +{ + current_draw_call_count_++; + vertex_submitted_count_ += vertex_submission; + empty_ = false; +} + +/* Reset workload counters. */ +void MTLCommandBufferManager::reset_counters() +{ + empty_ = true; + current_draw_call_count_ = 0; + encoder_count_ = 0; + vertex_submitted_count_ = 0; +} + +/* Workload evaluation. */ +bool MTLCommandBufferManager::do_break_submission() +{ + /* Skip if no active command buffer. */ + if (active_command_buffer_ == nil) { + return false; + } + + /* Use optimised heuristic to split heavy command buffer submissions to better saturate the + * hardware and also reduce stalling from individual large submissions. */ + if (GPU_type_matches(GPU_DEVICE_INTEL, GPU_OS_ANY, GPU_DRIVER_ANY) || + GPU_type_matches(GPU_DEVICE_ATI, GPU_OS_ANY, GPU_DRIVER_ANY)) { + return ((current_draw_call_count_ > 30000) || (vertex_submitted_count_ > 100000000) || + (encoder_count_ > 25)); + } + else { + /* Apple Silicon is less efficient if splitting submissions. */ + return false; + } +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \name Command buffer debugging. + * \{ */ + +/* Debug. */ +void MTLCommandBufferManager::push_debug_group(const char *name, int index) +{ + id<MTLCommandBuffer> cmd = this->ensure_begin(); + if (cmd != nil) { + [cmd pushDebugGroup:[NSString stringWithFormat:@"%s_%d", name, index]]; + } +} + +void MTLCommandBufferManager::pop_debug_group() +{ + id<MTLCommandBuffer> cmd = this->ensure_begin(); + if (cmd != nil) { + [cmd popDebugGroup]; + } +} + +/* Workload Synchronisation. */ +bool MTLCommandBufferManager::insert_memory_barrier(eGPUBarrier barrier_bits, + eGPUStageBarrierBits before_stages, + eGPUStageBarrierBits after_stages) +{ + /* Only supporting Metal on 10.14 onwards anyway - Check required for warnings. */ + if (@available(macOS 10.14, *)) { + + /* Resolve scope. */ + MTLBarrierScope scope = 0; + if (barrier_bits & GPU_BARRIER_SHADER_IMAGE_ACCESS || + barrier_bits & GPU_BARRIER_TEXTURE_FETCH) { + scope = scope | MTLBarrierScopeTextures | MTLBarrierScopeRenderTargets; + } + if (barrier_bits & GPU_BARRIER_SHADER_STORAGE || + barrier_bits & GPU_BARRIER_VERTEX_ATTRIB_ARRAY || + barrier_bits & GPU_BARRIER_ELEMENT_ARRAY) { + scope = scope | MTLBarrierScopeBuffers; + } + + if (scope != 0) { + /* Issue barrier based on encoder. */ + switch (active_command_encoder_type_) { + case MTL_NO_COMMAND_ENCODER: + case MTL_BLIT_COMMAND_ENCODER: { + /* No barrier to be inserted. */ + return false; + } + + /* Rendering. */ + case MTL_RENDER_COMMAND_ENCODER: { + /* Currently flagging both stages -- can use bits above to filter on stage type -- + * though full barrier is safe for now*/ + MTLRenderStages before_stage_flags = 0; + MTLRenderStages after_stage_flags = 0; + if (before_stages & GPU_BARRIER_STAGE_VERTEX && + !(before_stages & GPU_BARRIER_STAGE_FRAGMENT)) { + before_stage_flags = before_stage_flags | MTLRenderStageVertex; + } + if (before_stages & GPU_BARRIER_STAGE_FRAGMENT) { + before_stage_flags = before_stage_flags | MTLRenderStageFragment; + } + if (after_stages & GPU_BARRIER_STAGE_VERTEX) { + after_stage_flags = after_stage_flags | MTLRenderStageVertex; + } + if (after_stages & GPU_BARRIER_STAGE_FRAGMENT) { + after_stage_flags = MTLRenderStageFragment; + } + + id<MTLRenderCommandEncoder> rec = this->get_active_render_command_encoder(); + BLI_assert(rec != nil); + [rec memoryBarrierWithScope:scope + afterStages:after_stage_flags + beforeStages:before_stage_flags]; + return true; + } + + /* Compute. */ + case MTL_COMPUTE_COMMAND_ENCODER: { + id<MTLComputeCommandEncoder> rec = this->get_active_compute_command_encoder(); + BLI_assert(rec != nil); + [rec memoryBarrierWithScope:scope]; + return true; + } + } + } + } + /* No barrier support. */ + return false; +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \name Render Pass State for active RenderCommandEncoder + * \{ */ + +/* Metal Render Pass State. */ +void MTLRenderPassState::prepare(MTLCommandBufferManager *cmd, MTLContext *mtl_ctx) +{ + this->cmd = cmd; + this->ctx = mtl_ctx; + this->reset_state(); +} + +/* Reset binding state when a new RenderCommandEncoder is bound, to ensure + * pipeline resources are re-applied to the new Encoder. + * Note: In Metal, state is only persistent within an MTLCommandEncoder, + * not globally. */ +void MTLRenderPassState::reset_state() +{ + /* Reset Cached pipeline state. */ + this->bound_pso = nil; + this->bound_ds_state = nil; + + /* Clear shader binding. */ + this->last_bound_shader_state.set(nullptr, 0); + + /* Other states. */ + MTLFrameBuffer *fb = this->cmd->get_active_framebuffer(); + this->last_used_stencil_ref_value = 0; + this->last_scissor_rect = {0, + 0, + (unsigned long)((fb != nullptr) ? fb->get_width() : 0), + (unsigned long)((fb != nullptr) ? fb->get_height() : 0)}; + + /* Reset cached resource binding state */ + for (int ubo = 0; ubo < MTL_MAX_UNIFORM_BUFFER_BINDINGS; ubo++) { + this->cached_vertex_buffer_bindings[ubo].is_bytes = false; + this->cached_vertex_buffer_bindings[ubo].metal_buffer = nil; + this->cached_vertex_buffer_bindings[ubo].offset = -1; + + this->cached_fragment_buffer_bindings[ubo].is_bytes = false; + this->cached_fragment_buffer_bindings[ubo].metal_buffer = nil; + this->cached_fragment_buffer_bindings[ubo].offset = -1; + } + + /* Reset cached texture and sampler state binding state. */ + for (int tex = 0; tex < MTL_MAX_TEXTURE_SLOTS; tex++) { + this->cached_vertex_texture_bindings[tex].metal_texture = nil; + this->cached_vertex_sampler_state_bindings[tex].sampler_state = nil; + this->cached_vertex_sampler_state_bindings[tex].is_arg_buffer_binding = false; + + this->cached_fragment_texture_bindings[tex].metal_texture = nil; + this->cached_fragment_sampler_state_bindings[tex].sampler_state = nil; + this->cached_fragment_sampler_state_bindings[tex].is_arg_buffer_binding = false; + } +} + +/* Bind Texture to current RenderCommandEncoder. */ +void MTLRenderPassState::bind_vertex_texture(id<MTLTexture> tex, uint slot) +{ + if (this->cached_vertex_texture_bindings[slot].metal_texture != tex) { + id<MTLRenderCommandEncoder> rec = this->cmd->get_active_render_command_encoder(); + BLI_assert(rec != nil); + [rec setVertexTexture:tex atIndex:slot]; + this->cached_vertex_texture_bindings[slot].metal_texture = tex; + } +} + +void MTLRenderPassState::bind_fragment_texture(id<MTLTexture> tex, uint slot) +{ + if (this->cached_fragment_texture_bindings[slot].metal_texture != tex) { + id<MTLRenderCommandEncoder> rec = this->cmd->get_active_render_command_encoder(); + BLI_assert(rec != nil); + [rec setFragmentTexture:tex atIndex:slot]; + this->cached_fragment_texture_bindings[slot].metal_texture = tex; + } +} + +void MTLRenderPassState::bind_vertex_sampler(MTLSamplerBinding &sampler_binding, + bool use_argument_buffer_for_samplers, + uint slot) +{ + /* TODO(Metal): Implement RenderCommandEncoder vertex sampler binding utility. This will be + * implemented alongside MTLShader. */ +} + +void MTLRenderPassState::bind_fragment_sampler(MTLSamplerBinding &sampler_binding, + bool use_argument_buffer_for_samplers, + uint slot) +{ + /* TODO(Metal): Implement RenderCommandEncoder fragment sampler binding utility. This will be + * implemented alongside MTLShader. */ +} + +void MTLRenderPassState::bind_vertex_buffer(id<MTLBuffer> buffer, uint buffer_offset, uint index) +{ + /* TODO(Metal): Implement RenderCommandEncoder vertex buffer binding utility. This will be + * implemented alongside the full MTLMemoryManager. */ +} + +void MTLRenderPassState::bind_fragment_buffer(id<MTLBuffer> buffer, uint buffer_offset, uint index) +{ + /* TODO(Metal): Implement RenderCommandEncoder fragment buffer binding utility. This will be + * implemented alongside the full MTLMemoryManager. */ +} + +void MTLRenderPassState::bind_vertex_bytes(void *bytes, uint length, uint index) +{ + /* TODO(Metal): Implement RenderCommandEncoder vertex bytes binding utility. This will be + * implemented alongside the full MTLMemoryManager. */ +} + +void MTLRenderPassState::bind_fragment_bytes(void *bytes, uint length, uint index) +{ + /* TODO(Metal): Implement RenderCommandEncoder fragment bytes binding utility. This will be + * implemented alongside the full MTLMemoryManager. */ +} + +/** \} */ + +} // blender::gpu diff --git a/source/blender/gpu/metal/mtl_common.hh b/source/blender/gpu/metal/mtl_common.hh index aa60d3aff61..8dda2c43585 100644 --- a/source/blender/gpu/metal/mtl_common.hh +++ b/source/blender/gpu/metal/mtl_common.hh @@ -6,5 +6,6 @@ // -- Renderer Options -- #define MTL_MAX_SET_BYTES_SIZE 4096 #define MTL_FORCE_WAIT_IDLE 0 +#define MTL_MAX_COMMAND_BUFFERS 64 #endif diff --git a/source/blender/gpu/metal/mtl_context.hh b/source/blender/gpu/metal/mtl_context.hh index 1849a04ea48..92276a34b63 100644 --- a/source/blender/gpu/metal/mtl_context.hh +++ b/source/blender/gpu/metal/mtl_context.hh @@ -10,7 +10,9 @@ #include "GPU_common_types.h" #include "GPU_context.h" +#include "mtl_backend.hh" #include "mtl_capabilities.hh" +#include "mtl_framebuffer.hh" #include "mtl_texture.hh" #include <Cocoa/Cocoa.h> @@ -23,12 +25,118 @@ namespace blender::gpu { +/* Forward Declarations */ +class MTLContext; +class MTLCommandBufferManager; class MTLShader; class MTLUniformBuf; class MTLBuffer; +/* Structs containing information on current binding state for textures and samplers. */ +struct MTLTextureBinding { + bool used; + + /* Same value as index in bindings array. */ + uint texture_slot_index; + gpu::MTLTexture *texture_resource; +}; + +struct MTLSamplerBinding { + bool used; + MTLSamplerState state; + + bool operator==(MTLSamplerBinding const &other) const + { + return (used == other.used && state == other.state); + } +}; + +/* Metal Context Render Pass State -- Used to track active RenderCommandEncoder state based on + * bound MTLFrameBuffer's.Owned by MTLContext. */ +struct MTLRenderPassState { + friend class MTLContext; + + /* Given a RenderPassState is associated with a live RenderCommandEncoder, + * this state sits within the MTLCommandBufferManager. */ + MTLCommandBufferManager *cmd; + MTLContext *ctx; + + /* Caching of resource bindings for active MTLRenderCommandEncoder. + * In Metal, resource bindings are local to the MTLCommandEncoder, + * not globally to the whole pipeline/cmd buffer. */ + struct MTLBoundShaderState { + MTLShader *shader_ = nullptr; + uint pso_index_; + void set(MTLShader *shader, uint pso_index) + { + shader_ = shader; + pso_index_ = pso_index; + } + }; + + MTLBoundShaderState last_bound_shader_state; + id<MTLRenderPipelineState> bound_pso = nil; + id<MTLDepthStencilState> bound_ds_state = nil; + uint last_used_stencil_ref_value = 0; + MTLScissorRect last_scissor_rect; + + /* Caching of CommandEncoder Vertex/Fragment buffer bindings. */ + struct BufferBindingCached { + /* Whether the given binding slot uses byte data (Push Constant equivalent) + * or an MTLBuffer. */ + bool is_bytes; + id<MTLBuffer> metal_buffer; + int offset; + }; + + BufferBindingCached cached_vertex_buffer_bindings[MTL_MAX_UNIFORM_BUFFER_BINDINGS]; + BufferBindingCached cached_fragment_buffer_bindings[MTL_MAX_UNIFORM_BUFFER_BINDINGS]; + + /* Caching of CommandEncoder textures bindings. */ + struct TextureBindingCached { + id<MTLTexture> metal_texture; + }; + + TextureBindingCached cached_vertex_texture_bindings[MTL_MAX_TEXTURE_SLOTS]; + TextureBindingCached cached_fragment_texture_bindings[MTL_MAX_TEXTURE_SLOTS]; + + /* Cached of CommandEncoder sampler states. */ + struct SamplerStateBindingCached { + MTLSamplerState binding_state; + id<MTLSamplerState> sampler_state; + bool is_arg_buffer_binding; + }; + + SamplerStateBindingCached cached_vertex_sampler_state_bindings[MTL_MAX_TEXTURE_SLOTS]; + SamplerStateBindingCached cached_fragment_sampler_state_bindings[MTL_MAX_TEXTURE_SLOTS]; + + /* Prepare. */ + void prepare(MTLCommandBufferManager *cmd, MTLContext *ctx); + + /* Reset RenderCommandEncoder binding state. */ + void reset_state(); + + /* Texture Binding (RenderCommandEncoder). */ + void bind_vertex_texture(id<MTLTexture> tex, uint slot); + void bind_fragment_texture(id<MTLTexture> tex, uint slot); + + /* Sampler Binding (RenderCommandEncoder). */ + void bind_vertex_sampler(MTLSamplerBinding &sampler_binding, + bool use_argument_buffer_for_samplers, + uint slot); + void bind_fragment_sampler(MTLSamplerBinding &sampler_binding, + bool use_argument_buffer_for_samplers, + uint slot); + + /* Buffer binding (RenderCommandEncoder). */ + void bind_vertex_buffer(id<MTLBuffer> buffer, uint buffer_offset, uint index); + void bind_fragment_buffer(id<MTLBuffer> buffer, uint buffer_offset, uint index); + void bind_vertex_bytes(void *bytes, uint length, uint index); + void bind_fragment_bytes(void *bytes, uint length, uint index); +}; + /* Depth Stencil State */ -typedef struct MTLContextDepthStencilState { +struct MTLContextDepthStencilState { /* Depth State. */ bool depth_write_enable; @@ -44,9 +152,9 @@ typedef struct MTLContextDepthStencilState { /* Stencil State. */ bool stencil_test_enabled; - unsigned int stencil_read_mask; - unsigned int stencil_write_mask; - unsigned int stencil_ref; + uint stencil_read_mask; + uint stencil_write_mask; + uint stencil_ref; MTLCompareFunction stencil_func; MTLStencilOperation stencil_op_front_stencil_fail; @@ -65,7 +173,7 @@ typedef struct MTLContextDepthStencilState { /* TODO(Metal): Consider optimizing this function using memcmp. * Un-used, but differing, stencil state leads to over-generation * of state objects when doing trivial compare. */ - inline bool operator==(const MTLContextDepthStencilState &other) const + bool operator==(const MTLContextDepthStencilState &other) const { bool depth_state_equality = (has_depth_target == other.has_depth_target && depth_write_enable == other.depth_write_enable && @@ -98,7 +206,7 @@ typedef struct MTLContextDepthStencilState { * - setStencilReferenceValue: * - setDepthBias:slopeScale:clamp: */ - inline std::size_t hash() const + std::size_t hash() const { std::size_t boolean_bitmask = (this->depth_write_enable ? 1 : 0) | ((this->depth_test_enabled ? 1 : 0) << 1) | @@ -127,9 +235,9 @@ typedef struct MTLContextDepthStencilState { std::size_t final_hash = (main_hash << 8) | boolean_bitmask; return final_hash; } -} MTLContextDepthStencilState; +}; -typedef struct MTLContextTextureUtils { +struct MTLContextTextureUtils { /* Depth Update Utilities */ /* Depth texture updates are not directly supported with Blit operations, similarly, we cannot @@ -174,8 +282,7 @@ typedef struct MTLContextTextureUtils { blender::Map<TextureUpdateRoutineSpecialisation, id<MTLComputePipelineState>> texture_buffer_update_compute_psos; - template<typename T> - inline void free_cached_pso_map(blender::Map<T, id<MTLComputePipelineState>> &map) + template<typename T> void free_cached_pso_map(blender::Map<T, id<MTLComputePipelineState>> &map) { for (typename blender::Map<T, id<MTLComputePipelineState>>::MutableItem item : map.items()) { [item.value release]; @@ -183,12 +290,12 @@ typedef struct MTLContextTextureUtils { map.clear(); } - inline void init() + void init() { fullscreen_blit_shader = nullptr; } - inline void cleanup() + void cleanup() { if (fullscreen_blit_shader) { GPU_shader_free(fullscreen_blit_shader); @@ -213,37 +320,16 @@ typedef struct MTLContextTextureUtils { free_cached_pso_map(texture_cube_array_update_compute_psos); free_cached_pso_map(texture_buffer_update_compute_psos); } - -} MTLContextTextureUtils; - -/* Structs containing information on current binding state for textures and samplers. */ -typedef struct MTLTextureBinding { - bool used; - - /* Same value as index in bindings array. */ - unsigned int texture_slot_index; - gpu::MTLTexture *texture_resource; - -} MTLTextureBinding; - -typedef struct MTLSamplerBinding { - bool used; - MTLSamplerState state; - - bool operator==(MTLSamplerBinding const &other) const - { - return (used == other.used && state == other.state); - } -} MTLSamplerBinding; +}; /* Combined sampler state configuration for Argument Buffer caching. */ struct MTLSamplerArray { - unsigned int num_samplers; + uint num_samplers; /* MTLSamplerState permutations between 0..256 - slightly more than a byte. */ MTLSamplerState mtl_sampler_flags[MTL_MAX_TEXTURE_SLOTS]; id<MTLSamplerState> mtl_sampler[MTL_MAX_TEXTURE_SLOTS]; - inline bool operator==(const MTLSamplerArray &other) const + bool operator==(const MTLSamplerArray &other) const { if (this->num_samplers != other.num_samplers) { return false; @@ -253,7 +339,7 @@ struct MTLSamplerArray { sizeof(MTLSamplerState) * this->num_samplers) == 0); } - inline uint32_t hash() const + uint32_t hash() const { uint32_t hash = this->num_samplers; for (int i = 0; i < this->num_samplers; i++) { @@ -287,12 +373,12 @@ typedef enum MTLPipelineStateDirtyFlag { /* Ignore full flag bit-mask `MTL_PIPELINE_STATE_ALL_FLAG`. */ ENUM_OPERATORS(MTLPipelineStateDirtyFlag, MTL_PIPELINE_STATE_CULLMODE_FLAG); -typedef struct MTLUniformBufferBinding { +struct MTLUniformBufferBinding { bool bound; MTLUniformBuf *ubo; -} MTLUniformBufferBinding; +}; -typedef struct MTLContextGlobalShaderPipelineState { +struct MTLContextGlobalShaderPipelineState { bool initialised; /* Whether the pipeline state has been modified since application. @@ -358,11 +444,10 @@ typedef struct MTLContextGlobalShaderPipelineState { /* Render parameters. */ float point_size = 1.0f; float line_width = 1.0f; - -} MTLContextGlobalShaderPipelineState; +}; /* Metal Buffer */ -typedef struct MTLTemporaryBufferRange { +struct MTLTemporaryBufferRange { id<MTLBuffer> metal_buffer; void *host_ptr; unsigned long long buffer_offset; @@ -371,7 +456,118 @@ typedef struct MTLTemporaryBufferRange { void flush(); bool requires_flush(); -} MTLTemporaryBufferRange; +}; + +/* Command Buffer Manager - Owned by MTLContext. + * The MTLCommandBufferManager represents all work associated with + * a command buffer of a given identity. This manager is a fixed-state + * on the context, which coordinates the lifetime of command buffers + * for particular categories of work. + * + * This ensures operations on command buffers, and the state associated, + * is correctly tracked and managed. Workload submission and MTLCommandEncoder + * coordination is managed from here. + * + * There is currently only one MTLCommandBufferManager for managing submission + * of the "main" rendering commands. A secondary upload command buffer track, + * or asynchronous compute command buffer track may be added in the future. */ +class MTLCommandBufferManager { + friend class MTLContext; + + public: + /* Event to coordinate sequential execution across all "main" command buffers. */ + static id<MTLEvent> sync_event; + static unsigned long long event_signal_val; + + /* Counter for active command buffers. */ + static int num_active_cmd_bufs; + + private: + /* Associated Context and properties. */ + MTLContext *context_ = nullptr; + bool supports_render_ = false; + + /* CommandBuffer tracking. */ + id<MTLCommandBuffer> active_command_buffer_ = nil; + id<MTLCommandBuffer> last_submitted_command_buffer_ = nil; + + /* Active MTLCommandEncoders. */ + enum { + MTL_NO_COMMAND_ENCODER = 0, + MTL_RENDER_COMMAND_ENCODER = 1, + MTL_BLIT_COMMAND_ENCODER = 2, + MTL_COMPUTE_COMMAND_ENCODER = 3 + } active_command_encoder_type_ = MTL_NO_COMMAND_ENCODER; + + id<MTLRenderCommandEncoder> active_render_command_encoder_ = nil; + id<MTLBlitCommandEncoder> active_blit_command_encoder_ = nil; + id<MTLComputeCommandEncoder> active_compute_command_encoder_ = nil; + + /* State associated with active RenderCommandEncoder. */ + MTLRenderPassState render_pass_state_; + MTLFrameBuffer *active_frame_buffer_ = nullptr; + MTLRenderPassDescriptor *active_pass_descriptor_ = nullptr; + + /* Workload heuristics - We may need to split command buffers to optimise workload and balancing. + */ + int current_draw_call_count_ = 0; + int encoder_count_ = 0; + int vertex_submitted_count_ = 0; + bool empty_ = true; + + public: + void prepare(MTLContext *ctx, bool supports_render = true); + + /* If wait is true, CPU will stall until GPU work has completed. */ + bool submit(bool wait); + + /* Fetch/query current encoder. */ + bool is_inside_render_pass(); + bool is_inside_blit(); + bool is_inside_compute(); + id<MTLRenderCommandEncoder> get_active_render_command_encoder(); + id<MTLBlitCommandEncoder> get_active_blit_command_encoder(); + id<MTLComputeCommandEncoder> get_active_compute_command_encoder(); + MTLFrameBuffer *get_active_framebuffer(); + + /* RenderPassState for RenderCommandEncoder. */ + MTLRenderPassState &get_render_pass_state() + { + /* Render pass state should only be valid if we are inside a render pass. */ + BLI_assert(this->is_inside_render_pass()); + return render_pass_state_; + } + + /* Rendering Heuristics. */ + void register_draw_counters(int vertex_submission); + void reset_counters(); + bool do_break_submission(); + + /* Encoder and Pass management. */ + /* End currently active MTLCommandEncoder. */ + bool end_active_command_encoder(); + id<MTLRenderCommandEncoder> ensure_begin_render_command_encoder(MTLFrameBuffer *ctx_framebuffer, + bool force_begin, + bool *new_pass); + id<MTLBlitCommandEncoder> ensure_begin_blit_encoder(); + id<MTLComputeCommandEncoder> ensure_begin_compute_encoder(); + + /* Workload Synchronisation. */ + bool insert_memory_barrier(eGPUBarrier barrier_bits, + eGPUStageBarrierBits before_stages, + eGPUStageBarrierBits after_stages); + /* TODO(Metal): Support fences in command buffer class. */ + + /* Debug. */ + void push_debug_group(const char *name, int index); + void pop_debug_group(); + + private: + /* Begin new command buffer. */ + id<MTLCommandBuffer> ensure_begin(); + + void register_encoder_counters(); +}; /** MTLContext -- Core render loop and state management. **/ /* NOTE(Metal): Partial MTLContext stub to provide wrapper functionality @@ -397,6 +593,9 @@ class MTLContext : public Context { MTLSamplerArray samplers_; blender::Map<MTLSamplerArray, gpu::MTLBuffer *> cached_sampler_buffers_; + /* Frame. */ + bool is_inside_frame_ = false; + public: /* Shaders and Pipeline state. */ MTLContextGlobalShaderPipelineState pipeline_state; @@ -405,17 +604,22 @@ class MTLContext : public Context { id<MTLCommandQueue> queue = nil; id<MTLDevice> device = nil; + /* CommandBuffer managers. */ + MTLCommandBufferManager main_command_buffer; + /* GPUContext interface. */ MTLContext(void *ghost_window); ~MTLContext(); static void check_error(const char *info); - void activate(void) override; - void deactivate(void) override; + void activate() override; + void deactivate() override; + void begin_frame() override; + void end_frame() override; - void flush(void) override; - void finish(void) override; + void flush() override; + void finish() override; void memory_statistics_get(int *total_mem, int *free_mem) override; @@ -428,27 +632,32 @@ class MTLContext : public Context { * rendering, binding resources, setting global state, resource management etc; */ - /* Metal Context Core functions. */ - /* Command Buffer Management. */ - id<MTLCommandBuffer> get_active_command_buffer(); + /** Metal Context Core functions. **/ + + /* Bind framebuffer to context. */ + void framebuffer_bind(MTLFrameBuffer *framebuffer); - /* Render Pass State and Management. */ - void begin_render_pass(); - void end_render_pass(); - bool is_render_pass_active(); + /* Restore framebuffer used by active context to default backbuffer. */ + void framebuffer_restore(); - /* Texture Binding. */ - void texture_bind(gpu::MTLTexture *mtl_texture, unsigned int texture_unit); - void sampler_bind(MTLSamplerState, unsigned int sampler_unit); + /* Ensure a render-pass using the Context framebuffer (active_fb_) is in progress. */ + id<MTLRenderCommandEncoder> ensure_begin_render_pass(); + + MTLFrameBuffer *get_current_framebuffer(); + MTLFrameBuffer *get_default_framebuffer(); + + /* Context Global-State Texture Binding. */ + void texture_bind(gpu::MTLTexture *mtl_texture, uint texture_unit); + void sampler_bind(MTLSamplerState, uint sampler_unit); void texture_unbind(gpu::MTLTexture *mtl_texture); - void texture_unbind_all(void); + void texture_unbind_all(); id<MTLSamplerState> get_sampler_from_state(MTLSamplerState state); id<MTLSamplerState> generate_sampler_from_state(MTLSamplerState state); id<MTLSamplerState> get_default_sampler_state(); /* Metal Context pipeline state. */ - void pipeline_state_init(void); - MTLShader *get_active_shader(void); + void pipeline_state_init(); + MTLShader *get_active_shader(); /* State assignment. */ void set_viewport(int origin_x, int origin_y, int width, int height); @@ -458,7 +667,17 @@ class MTLContext : public Context { /* Texture utilities. */ MTLContextTextureUtils &get_texture_utils() { - return this->texture_utils_; + return texture_utils_; + } + + bool get_active() + { + return is_active_; + } + + bool get_inside_frame() + { + return is_inside_frame_; } }; diff --git a/source/blender/gpu/metal/mtl_context.mm b/source/blender/gpu/metal/mtl_context.mm index 94f5682b11b..64a44c6f718 100644 --- a/source/blender/gpu/metal/mtl_context.mm +++ b/source/blender/gpu/metal/mtl_context.mm @@ -22,7 +22,7 @@ namespace blender::gpu { bool MTLTemporaryBufferRange::requires_flush() { - /* We do not need to flush shared memory */ + /* We do not need to flush shared memory. */ return this->options & MTLResourceStorageModeManaged; } @@ -49,15 +49,86 @@ MTLContext::MTLContext(void *ghost_window) /* Init debug. */ debug::mtl_debug_init(); + /* Initialise command buffer state. */ + this->main_command_buffer.prepare(this); + + /* Frame management. */ + is_inside_frame_ = false; + + /* Create FrameBuffer handles. */ + MTLFrameBuffer *mtl_front_left = new MTLFrameBuffer(this, "front_left"); + MTLFrameBuffer *mtl_back_left = new MTLFrameBuffer(this, "back_left"); + this->front_left = mtl_front_left; + this->back_left = mtl_back_left; + this->active_fb = this->back_left; + /* Prepare platform and capabilities. (Note: With METAL, this needs to be done after CTX + * initialisation). */ + MTLBackend::platform_init(this); + MTLBackend::capabilities_init(this); /* Initialize Metal modules. */ this->state_manager = new MTLStateManager(this); - /* TODO(Metal): Implement. */ + /* Initialise texture read/update structures. */ + this->get_texture_utils().init(); + + /* Bound Samplers struct. */ + for (int i = 0; i < MTL_MAX_TEXTURE_SLOTS; i++) { + samplers_.mtl_sampler[i] = nil; + samplers_.mtl_sampler_flags[i] = DEFAULT_SAMPLER_STATE; + } + + /* Initialise samplers. */ + for (uint i = 0; i < GPU_SAMPLER_MAX; i++) { + MTLSamplerState state; + state.state = static_cast<eGPUSamplerState>(i); + sampler_state_cache_[i] = this->generate_sampler_from_state(state); + } } MTLContext::~MTLContext() { - /* TODO(Metal): Implement. */ + BLI_assert(this == reinterpret_cast<MTLContext *>(GPU_context_active_get())); + /* Ensure rendering is complete command encoders/command buffers are freed. */ + if (MTLBackend::get()->is_inside_render_boundary()) { + this->finish(); + + /* End frame. */ + if (is_inside_frame_) { + this->end_frame(); + } + } + /* Release update/blit shaders. */ + this->get_texture_utils().cleanup(); + + /* Release Sampler States. */ + for (int i = 0; i < GPU_SAMPLER_MAX; i++) { + if (sampler_state_cache_[i] != nil) { + [sampler_state_cache_[i] release]; + sampler_state_cache_[i] = nil; + } + } +} + +void MTLContext::begin_frame() +{ + BLI_assert(MTLBackend::get()->is_inside_render_boundary()); + if (is_inside_frame_) { + return; + } + + /* Begin Command buffer for next frame. */ + is_inside_frame_ = true; +} + +void MTLContext::end_frame() +{ + BLI_assert(is_inside_frame_); + + /* Ensure pre-present work is commited. */ + this->flush(); + + /* Increment frame counter. */ + is_inside_frame_ = false; } void MTLContext::check_error(const char *info) @@ -90,26 +161,83 @@ void MTLContext::memory_statistics_get(int *total_mem, int *free_mem) *free_mem = 0; } -id<MTLCommandBuffer> MTLContext::get_active_command_buffer() +void MTLContext::framebuffer_bind(MTLFrameBuffer *framebuffer) { - /* TODO(Metal): Implement. */ - return nil; + /* We do not yet begin the pass -- We defer beginning the pass until a draw is requested. */ + BLI_assert(framebuffer); + this->active_fb = framebuffer; } -/* Render Pass State and Management */ -void MTLContext::begin_render_pass() +void MTLContext::framebuffer_restore() { - /* TODO(Metal): Implement. */ + /* Bind default framebuffer from context -- + * We defer beginning the pass until a draw is requested. */ + this->active_fb = this->back_left; } -void MTLContext::end_render_pass() + +id<MTLRenderCommandEncoder> MTLContext::ensure_begin_render_pass() { - /* TODO(Metal): Implement. */ + BLI_assert(this); + + /* Ensure the rendering frame has started. */ + if (!is_inside_frame_) { + this->begin_frame(); + } + + /* Check whether a framebuffer is bound. */ + if (!this->active_fb) { + BLI_assert(false && "No framebuffer is bound!"); + return this->main_command_buffer.get_active_render_command_encoder(); + } + + /* Ensure command buffer workload submissions are optimal -- + * Though do not split a batch mid-IMM recording */ + /* TODO(Metal): Add IMM Check once MTLImmediate has been implemented. */ + if (this->main_command_buffer.do_break_submission()/*&& + !((MTLImmediate *)(this->imm))->imm_is_recording()*/) { + this->flush(); + } + + /* Begin pass or perform a pass switch if the active framebuffer has been changed, or if the + * framebuffer state has been modified (is_dirty). */ + if (!this->main_command_buffer.is_inside_render_pass() || + this->active_fb != this->main_command_buffer.get_active_framebuffer() || + this->main_command_buffer.get_active_framebuffer()->get_dirty()) { + + /* Validate bound framebuffer before beginning render pass. */ + if (!static_cast<MTLFrameBuffer *>(this->active_fb)->validate_render_pass()) { + MTL_LOG_WARNING("Framebuffer validation failed, falling back to default framebuffer\n"); + this->framebuffer_restore(); + + if (!static_cast<MTLFrameBuffer *>(this->active_fb)->validate_render_pass()) { + MTL_LOG_ERROR("CRITICAL: DEFAULT FRAMEBUFFER FAIL VALIDATION!!\n"); + } + } + + /* Begin RenderCommandEncoder on main CommandBuffer. */ + bool new_render_pass = false; + id<MTLRenderCommandEncoder> new_enc = + this->main_command_buffer.ensure_begin_render_command_encoder( + static_cast<MTLFrameBuffer *>(this->active_fb), true, &new_render_pass); + if (new_render_pass) { + /* Flag context pipeline state as dirty - dynamic pipeline state need re-applying. */ + this->pipeline_state.dirty_flags = MTL_PIPELINE_STATE_ALL_FLAG; + } + return new_enc; + } + BLI_assert(!this->main_command_buffer.get_active_framebuffer()->get_dirty()); + return this->main_command_buffer.get_active_render_command_encoder(); } -bool MTLContext::is_render_pass_active() +MTLFrameBuffer *MTLContext::get_current_framebuffer() { - /* TODO(Metal): Implement. */ - return false; + MTLFrameBuffer *last_bound = static_cast<MTLFrameBuffer *>(this->active_fb); + return last_bound ? last_bound : this->get_default_framebuffer(); +} + +MTLFrameBuffer *MTLContext::get_default_framebuffer() +{ + return static_cast<MTLFrameBuffer *>(this->back_left); } /** \} */ @@ -200,13 +328,68 @@ void MTLContext::pipeline_state_init() MTLStencilOperationKeep; } +void MTLContext::set_viewport(int origin_x, int origin_y, int width, int height) +{ + BLI_assert(this); + BLI_assert(width > 0); + BLI_assert(height > 0); + BLI_assert(origin_x >= 0); + BLI_assert(origin_y >= 0); + bool changed = (this->pipeline_state.viewport_offset_x != origin_x) || + (this->pipeline_state.viewport_offset_y != origin_y) || + (this->pipeline_state.viewport_width != width) || + (this->pipeline_state.viewport_height != height); + this->pipeline_state.viewport_offset_x = origin_x; + this->pipeline_state.viewport_offset_y = origin_y; + this->pipeline_state.viewport_width = width; + this->pipeline_state.viewport_height = height; + if (changed) { + this->pipeline_state.dirty_flags = (this->pipeline_state.dirty_flags | + MTL_PIPELINE_STATE_VIEWPORT_FLAG); + } +} + +void MTLContext::set_scissor(int scissor_x, int scissor_y, int scissor_width, int scissor_height) +{ + BLI_assert(this); + bool changed = (this->pipeline_state.scissor_x != scissor_x) || + (this->pipeline_state.scissor_y != scissor_y) || + (this->pipeline_state.scissor_width != scissor_width) || + (this->pipeline_state.scissor_height != scissor_height) || + (this->pipeline_state.scissor_enabled != true); + this->pipeline_state.scissor_x = scissor_x; + this->pipeline_state.scissor_y = scissor_y; + this->pipeline_state.scissor_width = scissor_width; + this->pipeline_state.scissor_height = scissor_height; + this->pipeline_state.scissor_enabled = (scissor_width > 0 && scissor_height > 0); + + if (changed) { + this->pipeline_state.dirty_flags = (this->pipeline_state.dirty_flags | + MTL_PIPELINE_STATE_SCISSOR_FLAG); + } +} + +void MTLContext::set_scissor_enabled(bool scissor_enabled) +{ + /* Only turn on Scissor if requested scissor region is valid */ + scissor_enabled = scissor_enabled && (this->pipeline_state.scissor_width > 0 && + this->pipeline_state.scissor_height > 0); + + bool changed = (this->pipeline_state.scissor_enabled != scissor_enabled); + this->pipeline_state.scissor_enabled = scissor_enabled; + if (changed) { + this->pipeline_state.dirty_flags = (this->pipeline_state.dirty_flags | + MTL_PIPELINE_STATE_SCISSOR_FLAG); + } +} + /** \} */ /* -------------------------------------------------------------------- */ /** \name Texture State Management * \{ */ -void MTLContext::texture_bind(gpu::MTLTexture *mtl_texture, unsigned int texture_unit) +void MTLContext::texture_bind(gpu::MTLTexture *mtl_texture, uint texture_unit) { BLI_assert(this); BLI_assert(mtl_texture); @@ -226,7 +409,7 @@ void MTLContext::texture_bind(gpu::MTLTexture *mtl_texture, unsigned int texture mtl_texture->is_bound_ = true; } -void MTLContext::sampler_bind(MTLSamplerState sampler_state, unsigned int sampler_unit) +void MTLContext::sampler_bind(MTLSamplerState sampler_state, uint sampler_unit) { BLI_assert(this); if (sampler_unit < 0 || sampler_unit >= GPU_max_textures() || @@ -271,14 +454,14 @@ void MTLContext::texture_unbind_all() id<MTLSamplerState> MTLContext::get_sampler_from_state(MTLSamplerState sampler_state) { - BLI_assert((unsigned int)sampler_state >= 0 && ((unsigned int)sampler_state) < GPU_SAMPLER_MAX); - return this->sampler_state_cache_[(unsigned int)sampler_state]; + BLI_assert((uint)sampler_state >= 0 && ((uint)sampler_state) < GPU_SAMPLER_MAX); + return sampler_state_cache_[(uint)sampler_state]; } id<MTLSamplerState> MTLContext::generate_sampler_from_state(MTLSamplerState sampler_state) { /* Check if sampler already exists for given state. */ - id<MTLSamplerState> st = this->sampler_state_cache_[(unsigned int)sampler_state]; + id<MTLSamplerState> st = sampler_state_cache_[(uint)sampler_state]; if (st != nil) { return st; } @@ -318,7 +501,7 @@ id<MTLSamplerState> MTLContext::generate_sampler_from_state(MTLSamplerState samp descriptor.supportArgumentBuffers = true; id<MTLSamplerState> state = [this->device newSamplerStateWithDescriptor:descriptor]; - this->sampler_state_cache_[(unsigned int)sampler_state] = state; + sampler_state_cache_[(uint)sampler_state] = state; BLI_assert(state != nil); [descriptor autorelease]; @@ -328,10 +511,10 @@ id<MTLSamplerState> MTLContext::generate_sampler_from_state(MTLSamplerState samp id<MTLSamplerState> MTLContext::get_default_sampler_state() { - if (this->default_sampler_state_ == nil) { - this->default_sampler_state_ = this->get_sampler_from_state(DEFAULT_SAMPLER_STATE); + if (default_sampler_state_ == nil) { + default_sampler_state_ = this->get_sampler_from_state(DEFAULT_SAMPLER_STATE); } - return this->default_sampler_state_; + return default_sampler_state_; } /** \} */ diff --git a/source/blender/gpu/metal/mtl_debug.mm b/source/blender/gpu/metal/mtl_debug.mm index 9d67a1f4f04..8ca4a0cc6e3 100644 --- a/source/blender/gpu/metal/mtl_debug.mm +++ b/source/blender/gpu/metal/mtl_debug.mm @@ -46,20 +46,14 @@ namespace blender::gpu { void MTLContext::debug_group_begin(const char *name, int index) { if (G.debug & G_DEBUG_GPU) { - id<MTLCommandBuffer> cmd = this->get_active_command_buffer(); - if (cmd != nil) { - [cmd pushDebugGroup:[NSString stringWithFormat:@"%s_%d", name, index]]; - } + this->main_command_buffer.push_debug_group(name, index); } } void MTLContext::debug_group_end() { if (G.debug & G_DEBUG_GPU) { - id<MTLCommandBuffer> cmd = this->get_active_command_buffer(); - if (cmd != nil) { - [cmd popDebugGroup]; - } + this->main_command_buffer.pop_debug_group(); } } diff --git a/source/blender/gpu/metal/mtl_framebuffer.hh b/source/blender/gpu/metal/mtl_framebuffer.hh new file mode 100644 index 00000000000..6849e574d81 --- /dev/null +++ b/source/blender/gpu/metal/mtl_framebuffer.hh @@ -0,0 +1,231 @@ +/** \file + * \ingroup gpu + * + * Encapsulation of Framebuffer states (attached textures, viewport, scissors). + */ + +#pragma once + +#include "GPU_common_types.h" +#include "MEM_guardedalloc.h" + +#include "gpu_framebuffer_private.hh" +#include "mtl_texture.hh" +#include <Metal/Metal.h> + +namespace blender::gpu { + +class MTLContext; + +struct MTLAttachment { + bool used; + gpu::MTLTexture *texture; + union { + float color[4]; + float depth; + uint stencil; + } clear_value; + + eGPULoadOp load_action; + eGPUStoreOp store_action; + uint mip; + uint slice; + uint depth_plane; + + /* If Array Length is larger than zero, use multilayered rendering. */ + uint render_target_array_length; +}; + +/** + * Implementation of FrameBuffer object using Metal. + **/ +class MTLFrameBuffer : public FrameBuffer { + private: + /* Context Handle. */ + MTLContext *context_; + + /* Metal Attachment properties. */ + uint colour_attachment_count_; + MTLAttachment mtl_color_attachments_[GPU_FB_MAX_COLOR_ATTACHMENT]; + MTLAttachment mtl_depth_attachment_; + MTLAttachment mtl_stencil_attachment_; + bool use_multilayered_rendering_ = false; + + /* State. */ + /* Whether global framebuffer properties have changed and require + * re-generation of MTLRenderPassDescriptor/RenderCommandEncoders. */ + bool is_dirty_; + + /* Whether loadstore properties have changed (only affects certain cached configs). */ + bool is_loadstore_dirty_; + + /* Context that the latest modified state was last applied to. + * If this does not match current ctx, re-apply state. */ + MTLContext *dirty_state_ctx_; + + /* Whether a clear is pending -- Used to toggle between clear and load FB configurations + * (without dirtying the state) - Framebuffer load config is used if no GPU_clear_* command + * was issued after binding the FrameBuffer. */ + bool has_pending_clear_; + + /* Render Pass Descriptors: + * There are 3 MTLRenderPassDescriptors for different ways in which a framebuffer + * can be configured: + * [0] = CLEAR CONFIG -- Used when a GPU_framebuffer_clear_* command has been issued. + * [1] = LOAD CONFIG -- Used if bound, but no clear is required. + * [2] = CUSTOM CONFIG -- When using GPU_framebuffer_bind_ex to manually specify + * load-store configuration for optimal bandwidth utilisation. + * -- We cache these different configs to avoid re-generation -- + */ + typedef enum { + MTL_FB_CONFIG_CLEAR = 0, + MTL_FB_CONFIG_LOAD = 1, + MTL_FB_CONFIG_CUSTOM = 2 + } MTL_FB_CONFIG; +#define MTL_FB_CONFIG_MAX (MTL_FB_CONFIG_CUSTOM + 1) + + MTLRenderPassDescriptor *framebuffer_descriptor_[MTL_FB_CONFIG_MAX]; + MTLRenderPassColorAttachmentDescriptor + *colour_attachment_descriptors_[GPU_FB_MAX_COLOR_ATTACHMENT]; + /* Whether MTLRenderPassDescriptor[N] requires updating with latest state. */ + bool descriptor_dirty_[MTL_FB_CONFIG_MAX]; + /* Whether SRGB is enabled for this framebuffer configuration. */ + bool srgb_enabled_; + /* Whether the primary Framebuffer attachment is an SRGB target or not. */ + bool is_srgb_; + + public: + /** + * Create a conventional framebuffer to attach texture to. + **/ + MTLFrameBuffer(MTLContext *ctx, const char *name); + + ~MTLFrameBuffer(); + + void bind(bool enabled_srgb) override; + + bool check(char err_out[256]) override; + + void clear(eGPUFrameBufferBits buffers, + const float clear_col[4], + float clear_depth, + uint clear_stencil) override; + void clear_multi(const float (*clear_cols)[4]) override; + void clear_attachment(GPUAttachmentType type, + eGPUDataFormat data_format, + const void *clear_value) override; + + void attachment_set_loadstore_op(GPUAttachmentType type, + eGPULoadOp load_action, + eGPUStoreOp store_action) override; + + void read(eGPUFrameBufferBits planes, + eGPUDataFormat format, + const int area[4], + int channel_len, + int slot, + void *r_data) override; + + void blit_to(eGPUFrameBufferBits planes, + int src_slot, + FrameBuffer *dst, + int dst_slot, + int dst_offset_x, + int dst_offset_y) override; + + void apply_state(); + + /* State. */ + /* Flag MTLFramebuffer configuration as having changed. */ + void mark_dirty(); + void mark_loadstore_dirty(); + /* Mark that a pending clear has been performed. */ + void mark_cleared(); + /* Mark that we have a pending clear. */ + void mark_do_clear(); + + /* Attachment management. */ + /* When dirty_attachments_ is true, we need to reprocess attachments to extract Metal + * information. */ + void update_attachments(bool update_viewport); + bool add_color_attachment(gpu::MTLTexture *texture, uint slot, int miplevel, int layer); + bool add_depth_attachment(gpu::MTLTexture *texture, int miplevel, int layer); + bool add_stencil_attachment(gpu::MTLTexture *texture, int miplevel, int layer); + bool remove_color_attachment(uint slot); + bool remove_depth_attachment(); + bool remove_stencil_attachment(); + void remove_all_attachments(); + void ensure_render_target_size(); + + /* Clear values -> Load/store actions. */ + bool set_color_attachment_clear_color(uint slot, const float clear_color[4]); + bool set_depth_attachment_clear_value(float depth_clear); + bool set_stencil_attachment_clear_value(uint stencil_clear); + bool set_color_loadstore_op(uint slot, eGPULoadOp load_action, eGPUStoreOp store_action); + bool set_depth_loadstore_op(eGPULoadOp load_action, eGPUStoreOp store_action); + bool set_stencil_loadstore_op(eGPULoadOp load_action, eGPUStoreOp store_action); + + /* Remove any pending clears - Ensure "load" configuration is used. */ + bool reset_clear_state(); + + /* Fetch values */ + bool has_attachment_at_slot(uint slot); + bool has_color_attachment_with_texture(gpu::MTLTexture *texture); + bool has_depth_attachment(); + bool has_stencil_attachment(); + int get_color_attachment_slot_from_texture(gpu::MTLTexture *texture); + uint get_attachment_count(); + uint get_attachment_limit() + { + return GPU_FB_MAX_COLOR_ATTACHMENT; + }; + MTLAttachment get_color_attachment(uint slot); + MTLAttachment get_depth_attachment(); + MTLAttachment get_stencil_attachment(); + + /* Metal API resources and validation. */ + bool validate_render_pass(); + MTLRenderPassDescriptor *bake_render_pass_descriptor(bool load_contents); + + /* Blitting. */ + void blit(uint read_slot, + uint src_x_offset, + uint src_y_offset, + MTLFrameBuffer *metal_fb_write, + uint write_slot, + uint dst_x_offset, + uint dst_y_offset, + uint width, + uint height, + eGPUFrameBufferBits blit_buffers); + + int get_width(); + int get_height(); + bool get_dirty() + { + return is_dirty_ || is_loadstore_dirty_; + } + + bool get_pending_clear() + { + return has_pending_clear_; + } + + bool get_srgb_enabled() + { + return srgb_enabled_; + } + + bool get_is_srgb() + { + return is_srgb_; + } + + private: + /* Clears a render target by force-opening a render pass. */ + void force_clear(); + + MEM_CXX_CLASS_ALLOC_FUNCS("MTLFrameBuffer"); +}; + +} // namespace blender::gpu diff --git a/source/blender/gpu/metal/mtl_framebuffer.mm b/source/blender/gpu/metal/mtl_framebuffer.mm new file mode 100644 index 00000000000..b2cab3f8af6 --- /dev/null +++ b/source/blender/gpu/metal/mtl_framebuffer.mm @@ -0,0 +1,1897 @@ +/** \file + * \ingroup gpu + */ + +#include "BKE_global.h" + +#include "mtl_context.hh" +#include "mtl_debug.hh" +#include "mtl_framebuffer.hh" +#include "mtl_texture.hh" +#import <Availability.h> + +namespace blender::gpu { + +/* -------------------------------------------------------------------- */ +/** \name Creation & Deletion + * \{ */ + +MTLFrameBuffer::MTLFrameBuffer(MTLContext *ctx, const char *name) : FrameBuffer(name) +{ + + context_ = ctx; + is_dirty_ = true; + is_loadstore_dirty_ = true; + dirty_state_ctx_ = nullptr; + has_pending_clear_ = false; + colour_attachment_count_ = 0; + srgb_enabled_ = false; + is_srgb_ = false; + + for (int i = 0; i < GPU_FB_MAX_COLOR_ATTACHMENT; i++) { + mtl_color_attachments_[i].used = false; + } + mtl_depth_attachment_.used = false; + mtl_stencil_attachment_.used = false; + + for (int i = 0; i < MTL_FB_CONFIG_MAX; i++) { + framebuffer_descriptor_[i] = [[MTLRenderPassDescriptor alloc] init]; + descriptor_dirty_[i] = true; + } + + for (int i = 0; i < GPU_FB_MAX_COLOR_ATTACHMENT; i++) { + colour_attachment_descriptors_[i] = [[MTLRenderPassColorAttachmentDescriptor alloc] init]; + } + + /* Initial state. */ + this->size_set(0, 0); + this->viewport_reset(); + this->scissor_reset(); +} + +MTLFrameBuffer::~MTLFrameBuffer() +{ + /* If FrameBuffer is associated with a currently open RenderPass, end. */ + if (context_->main_command_buffer.get_active_framebuffer() == this) { + context_->main_command_buffer.end_active_command_encoder(); + } + + /* Restore default frame-buffer if this frame-buffer was bound. */ + if (context_->active_fb == this && context_->back_left != this) { + /* If this assert triggers it means the frame-buffer is being freed while in use by another + * context which, by the way, is TOTALLY UNSAFE!!! (Copy from GL behaviour). */ + BLI_assert(context_ == static_cast<MTLContext *>(unwrap(GPU_context_active_get()))); + GPU_framebuffer_restore(); + } + + /* Free Render Pass Descriptors. */ + for (int config = 0; config < MTL_FB_CONFIG_MAX; config++) { + if (framebuffer_descriptor_[config] != nil) { + [framebuffer_descriptor_[config] release]; + framebuffer_descriptor_[config] = nil; + } + } + + /* Free colour attachment descriptors. */ + for (int i = 0; i < GPU_FB_MAX_COLOR_ATTACHMENT; i++) { + if (colour_attachment_descriptors_[i] != nil) { + [colour_attachment_descriptors_[i] release]; + colour_attachment_descriptors_[i] = nil; + } + } + + /* Remove attachments - release FB texture references. */ + this->remove_all_attachments(); + + if (context_ == nullptr) { + return; + } +} + +void MTLFrameBuffer::bind(bool enabled_srgb) +{ + + /* Verify Context is valid. */ + if (context_ != static_cast<MTLContext *>(unwrap(GPU_context_active_get()))) { + BLI_assert(false && "Trying to use the same frame-buffer in multiple context's."); + return; + } + + /* Ensure SRGB state is up-to-date and valid. */ + bool srgb_state_changed = srgb_enabled_ != enabled_srgb; + if (context_->active_fb != this || srgb_state_changed) { + if (srgb_state_changed) { + this->mark_dirty(); + } + srgb_enabled_ = enabled_srgb; + GPU_shader_set_framebuffer_srgb_target(srgb_enabled_ && is_srgb_); + } + + /* Ensure local MTLAttachment data is up to date. */ + this->update_attachments(true); + + /* Reset clear state on bind -- Clears and load/store ops are set after binding. */ + this->reset_clear_state(); + + /* Bind to active context. */ + MTLContext *mtl_context = reinterpret_cast<MTLContext *>(GPU_context_active_get()); + if (mtl_context) { + mtl_context->framebuffer_bind(this); + dirty_state_ = true; + } + else { + MTL_LOG_WARNING("Attempting to bind FrameBuffer, but no context is active\n"); + } +} + +bool MTLFrameBuffer::check(char err_out[256]) +{ + /* Ensure local MTLAttachment data is up to date. */ + this->update_attachments(true); + + /* Ensure there is atleast one attachment. */ + bool valid = (this->get_attachment_count() > 0 || + this->has_depth_attachment() | this->has_stencil_attachment()); + if (!valid) { + const char *format = "Framebuffer %s does not have any attachments.\n"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + MTL_LOG_ERROR(format, name_); + } + return false; + } + + /* Ensure all attachments have identical dimensions. */ + /* Ensure all attachments are rendertargets. */ + bool first = true; + uint dim_x = 0; + uint dim_y = 0; + for (int col_att = 0; col_att < this->get_attachment_count(); col_att++) { + MTLAttachment att = this->get_color_attachment(col_att); + if (att.used) { + if (att.texture->gpu_image_usage_flags_ & GPU_TEXTURE_USAGE_ATTACHMENT) { + if (first) { + dim_x = att.texture->width_get(); + dim_y = att.texture->height_get(); + first = false; + } + else { + if (dim_x != att.texture->width_get() || dim_y != att.texture->height_get()) { + const char *format = + "Framebuffer %s:Colour attachment dimensions do not match those of previous " + "attachment\n"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + fprintf(stderr, format, name_); + MTL_LOG_ERROR(format, name_); + } + return false; + } + } + } + else { + const char *format = + "Framebuffer %s: Colour attachment texture does not have usage flag " + "'GPU_TEXTURE_USAGE_ATTACHMENT'\n"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + fprintf(stderr, format, name_); + MTL_LOG_ERROR(format, name_); + } + return false; + } + } + } + MTLAttachment depth_att = this->get_depth_attachment(); + MTLAttachment stencil_att = this->get_stencil_attachment(); + if (depth_att.used) { + if (first) { + dim_x = depth_att.texture->width_get(); + dim_y = depth_att.texture->height_get(); + first = false; + valid = (depth_att.texture->gpu_image_usage_flags_ & GPU_TEXTURE_USAGE_ATTACHMENT); + + if (!valid) { + const char *format = + "Framebuffer %n: Depth attachment does not have usage " + "'GPU_TEXTURE_USAGE_ATTACHMENT'\n"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + fprintf(stderr, format, name_); + MTL_LOG_ERROR(format, name_); + } + return false; + } + } + else { + if (dim_x != depth_att.texture->width_get() || dim_y != depth_att.texture->height_get()) { + const char *format = + "Framebuffer %n: Depth attachment dimensions do not match that of previous " + "attachment\n"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + fprintf(stderr, format, name_); + MTL_LOG_ERROR(format, name_); + } + return false; + } + } + } + if (stencil_att.used) { + if (first) { + dim_x = stencil_att.texture->width_get(); + dim_y = stencil_att.texture->height_get(); + first = false; + valid = (stencil_att.texture->gpu_image_usage_flags_ & GPU_TEXTURE_USAGE_ATTACHMENT); + if (!valid) { + const char *format = + "Framebuffer %s: Stencil attachment does not have usage " + "'GPU_TEXTURE_USAGE_ATTACHMENT'\n"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + fprintf(stderr, format, name_); + MTL_LOG_ERROR(format, name_); + } + return false; + } + } + else { + if (dim_x != stencil_att.texture->width_get() || + dim_y != stencil_att.texture->height_get()) { + const char *format = + "Framebuffer %s: Stencil attachment dimensions do not match that of previous " + "attachment"; + if (err_out) { + BLI_snprintf(err_out, 256, format, name_); + } + else { + fprintf(stderr, format, name_); + MTL_LOG_ERROR(format, name_); + } + return false; + } + } + } + + BLI_assert(valid); + return valid; +} + +void MTLFrameBuffer::force_clear() +{ + /* Perform clear by ending current and starting a new render pass. */ + MTLContext *mtl_context = static_cast<MTLContext *>(unwrap(GPU_context_active_get())); + MTLFrameBuffer *current_framebuffer = mtl_context->get_current_framebuffer(); + if (current_framebuffer) { + BLI_assert(current_framebuffer == this); + /* End current renderpass. */ + if (mtl_context->main_command_buffer.is_inside_render_pass()) { + mtl_context->main_command_buffer.end_active_command_encoder(); + } + mtl_context->ensure_begin_render_pass(); + BLI_assert(has_pending_clear_ == false); + } +} + +void MTLFrameBuffer::clear(eGPUFrameBufferBits buffers, + const float clear_col[4], + float clear_depth, + uint clear_stencil) +{ + + BLI_assert(unwrap(GPU_context_active_get()) == context_); + BLI_assert(context_->active_fb == this); + + /* Ensure attachments are up to date. */ + this->update_attachments(true); + + /* If we had no previous clear pending, reset clear state. */ + if (!has_pending_clear_) { + this->reset_clear_state(); + } + + /* Ensure we only clear if attachments exist for given buffer bits. */ + bool do_clear = false; + if (buffers & GPU_COLOR_BIT) { + for (int i = 0; i < colour_attachment_count_; i++) { + this->set_color_attachment_clear_color(i, clear_col); + do_clear = true; + } + } + + if (buffers & GPU_DEPTH_BIT) { + this->set_depth_attachment_clear_value(clear_depth); + do_clear = do_clear || this->has_depth_attachment(); + } + if (buffers & GPU_STENCIL_BIT) { + this->set_stencil_attachment_clear_value(clear_stencil); + do_clear = do_clear || this->has_stencil_attachment(); + } + + if (do_clear) { + has_pending_clear_ = true; + + /* Apply state before clear. */ + this->apply_state(); + + /* TODO(Metal): Optimise - Currently force-clear always used. Consider moving clear state to + * MTLTexture instead. */ + /* Force clear if RP is not yet active -- not the most efficient, but there is no distinction + * between clears where no draws occur. Can optimise at the high-level by using explicit + * load-store flags. */ + this->force_clear(); + } +} + +void MTLFrameBuffer::clear_multi(const float (*clear_cols)[4]) +{ + /* If we had no previous clear pending, reset clear state. */ + if (!has_pending_clear_) { + this->reset_clear_state(); + } + + bool do_clear = false; + for (int i = 0; i < this->get_attachment_limit(); i++) { + if (this->has_attachment_at_slot(i)) { + this->set_color_attachment_clear_color(i, clear_cols[i]); + do_clear = true; + } + } + + if (do_clear) { + has_pending_clear_ = true; + + /* Apply state before clear. */ + this->apply_state(); + + /* TODO(Metal): Optimise - Currently force-clear always used. Consider moving clear state to + * MTLTexture instead. */ + /* Force clear if RP is not yet active -- not the most efficient, but there is no distinction + * between clears where no draws occur. Can optimise at the high-level by using explicit + * load-store flags. */ + this->force_clear(); + } +} + +void MTLFrameBuffer::clear_attachment(GPUAttachmentType type, + eGPUDataFormat data_format, + const void *clear_value) +{ + BLI_assert(static_cast<MTLContext *>(unwrap(GPU_context_active_get())) == context_); + BLI_assert(context_->active_fb == this); + + /* If we had no previous clear pending, reset clear state. */ + if (!has_pending_clear_) { + this->reset_clear_state(); + } + + bool do_clear = false; + + if (type == GPU_FB_DEPTH_STENCIL_ATTACHMENT) { + if (this->has_depth_attachment() || this->has_stencil_attachment()) { + BLI_assert(data_format == GPU_DATA_UINT_24_8); + float depth = ((*(uint32_t *)clear_value) & 0x00FFFFFFu) / (float)0x00FFFFFFu; + int stencil = ((*(uint32_t *)clear_value) >> 24); + this->set_depth_attachment_clear_value(depth); + this->set_stencil_attachment_clear_value(stencil); + do_clear = true; + } + } + else if (type == GPU_FB_DEPTH_ATTACHMENT) { + if (this->has_depth_attachment()) { + if (data_format == GPU_DATA_FLOAT) { + this->set_depth_attachment_clear_value(*(float *)clear_value); + } + else { + float depth = *(uint32_t *)clear_value / (float)0xFFFFFFFFu; + this->set_depth_attachment_clear_value(depth); + } + do_clear = true; + } + } + else { + int slot = type - GPU_FB_COLOR_ATTACHMENT0; + if (this->has_attachment_at_slot(slot)) { + float col_clear_val[4] = {0.0}; + switch (data_format) { + case GPU_DATA_FLOAT: { + const float *vals = (float *)clear_value; + col_clear_val[0] = vals[0]; + col_clear_val[1] = vals[1]; + col_clear_val[2] = vals[2]; + col_clear_val[3] = vals[3]; + } break; + case GPU_DATA_UINT: { + const uint *vals = (uint *)clear_value; + col_clear_val[0] = (float)(vals[0]); + col_clear_val[1] = (float)(vals[1]); + col_clear_val[2] = (float)(vals[2]); + col_clear_val[3] = (float)(vals[3]); + } break; + case GPU_DATA_INT: { + const int *vals = (int *)clear_value; + col_clear_val[0] = (float)(vals[0]); + col_clear_val[1] = (float)(vals[1]); + col_clear_val[2] = (float)(vals[2]); + col_clear_val[3] = (float)(vals[3]); + } break; + default: + BLI_assert_msg(0, "Unhandled data format"); + break; + } + this->set_color_attachment_clear_color(slot, col_clear_val); + do_clear = true; + } + } + + if (do_clear) { + has_pending_clear_ = true; + + /* Apply state before clear. */ + this->apply_state(); + + /* TODO(Metal): Optimise - Currently force-clear always used. Consider moving clear state to + * MTLTexture instead. */ + /* Force clear if RP is not yet active -- not the most efficient, but there is no distinction + * between clears where no draws occur. Can optimise at the high-level by using explicit + * load-store flags. */ + this->force_clear(); + } +} + +void MTLFrameBuffer::read(eGPUFrameBufferBits planes, + eGPUDataFormat format, + const int area[4], + int channel_len, + int slot, + void *r_data) +{ + + BLI_assert((planes & GPU_STENCIL_BIT) == 0); + BLI_assert(area[2] > 0); + BLI_assert(area[3] > 0); + + switch (planes) { + case GPU_DEPTH_BIT: { + if (this->has_depth_attachment()) { + MTLAttachment depth = this->get_depth_attachment(); + gpu::MTLTexture *tex = depth.texture; + if (tex) { + size_t sample_len = area[2] * area[3]; + size_t sample_size = to_bytesize(tex->format_, format); + int debug_data_size = sample_len * sample_size; + tex->read_internal(0, + area[0], + area[1], + 0, + area[2], + area[3], + 1, + format, + channel_len, + debug_data_size, + r_data); + } + } + else { + MTL_LOG_ERROR( + "Attempting to read depth from a framebuffer which does not have a depth " + "attachment!\n"); + } + } + return; + + case GPU_COLOR_BIT: { + if (this->has_attachment_at_slot(slot)) { + MTLAttachment color = this->get_color_attachment(slot); + gpu::MTLTexture *tex = color.texture; + if (tex) { + size_t sample_len = area[2] * area[3]; + size_t sample_size = to_bytesize(tex->format_, format); + int debug_data_size = sample_len * sample_size * channel_len; + tex->read_internal(0, + area[0], + area[1], + 0, + area[2], + area[3], + 1, + format, + channel_len, + debug_data_size, + r_data); + } + } + } + return; + + case GPU_STENCIL_BIT: + MTL_LOG_ERROR("GPUFramebuffer: Error: Trying to read stencil bit. Unsupported.\n"); + return; + } +} + +void MTLFrameBuffer::blit_to(eGPUFrameBufferBits planes, + int src_slot, + FrameBuffer *dst, + int dst_slot, + int dst_offset_x, + int dst_offset_y) +{ + this->update_attachments(true); + static_cast<MTLFrameBuffer *>(dst)->update_attachments(true); + + BLI_assert(planes != 0); + + MTLFrameBuffer *metal_fb_write = static_cast<MTLFrameBuffer *>(dst); + + BLI_assert(this); + BLI_assert(metal_fb_write); + + /* Get width/height from attachment. */ + MTLAttachment src_attachment; + const bool do_color = (planes & GPU_COLOR_BIT); + const bool do_depth = (planes & GPU_DEPTH_BIT); + const bool do_stencil = (planes & GPU_STENCIL_BIT); + + if (do_color) { + BLI_assert(!do_depth && !do_stencil); + src_attachment = this->get_color_attachment(src_slot); + } + else if (do_depth) { + BLI_assert(!do_color && !do_stencil); + src_attachment = this->get_depth_attachment(); + } + else if (do_stencil) { + BLI_assert(!do_color && !do_depth); + src_attachment = this->get_stencil_attachment(); + } + + BLI_assert(src_attachment.used); + this->blit(src_slot, + 0, + 0, + metal_fb_write, + dst_slot, + dst_offset_x, + dst_offset_y, + src_attachment.texture->width_get(), + src_attachment.texture->height_get(), + planes); +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \ Private METAL implementation functions + * \{ */ + +void MTLFrameBuffer::mark_dirty() +{ + is_dirty_ = true; + is_loadstore_dirty_ = true; +} + +void MTLFrameBuffer::mark_loadstore_dirty() +{ + is_loadstore_dirty_ = true; +} + +void MTLFrameBuffer::mark_cleared() +{ + has_pending_clear_ = false; +} + +void MTLFrameBuffer::mark_do_clear() +{ + has_pending_clear_ = true; +} + +void MTLFrameBuffer::update_attachments(bool update_viewport) +{ + if (!dirty_attachments_) { + return; + } + + /* Cache viewport and scissor (If we have existing attachments). */ + int t_viewport[4], t_scissor[4]; + update_viewport = update_viewport && + (this->get_attachment_count() > 0 && this->has_depth_attachment() && + this->has_stencil_attachment()); + if (update_viewport) { + this->viewport_get(t_viewport); + this->scissor_get(t_scissor); + } + + /* Clear current attachments state. */ + this->remove_all_attachments(); + + /* Reset framebuffer options. */ + use_multilayered_rendering_ = false; + + /* Track first attachment for SRGB property extraction. */ + GPUAttachmentType first_attachment = GPU_FB_MAX_ATTACHMENT; + MTLAttachment first_attachment_mtl; + + /* Scan through changes to attachments and populate local structures. */ + bool depth_added = false; + for (GPUAttachmentType type = GPU_FB_MAX_ATTACHMENT - 1; type >= 0; --type) { + GPUAttachment &attach = attachments_[type]; + + switch (type) { + case GPU_FB_DEPTH_ATTACHMENT: + case GPU_FB_DEPTH_STENCIL_ATTACHMENT: { + /* If one of the DEPTH types has added a texture, we avoid running this again, as it would + * only remove the target. */ + if (depth_added) { + break; + } + if (attach.tex) { + /* If we already had a depth attachment, preserve load/clear-state parameters, + * but remove existing and add new attachment. */ + if (this->has_depth_attachment()) { + MTLAttachment depth_attachment_prev = this->get_depth_attachment(); + this->remove_depth_attachment(); + this->add_depth_attachment( + static_cast<gpu::MTLTexture *>(unwrap(attach.tex)), attach.mip, attach.layer); + this->set_depth_attachment_clear_value(depth_attachment_prev.clear_value.depth); + this->set_depth_loadstore_op(depth_attachment_prev.load_action, + depth_attachment_prev.store_action); + } + else { + this->add_depth_attachment( + static_cast<gpu::MTLTexture *>(unwrap(attach.tex)), attach.mip, attach.layer); + } + + /* Check stencil component -- if supplied texture format supports stencil. */ + eGPUTextureFormat format = GPU_texture_format(attach.tex); + bool use_stencil = (type == GPU_FB_DEPTH_STENCIL_ATTACHMENT) && + (format == GPU_DEPTH32F_STENCIL8 || format == GPU_DEPTH24_STENCIL8); + if (use_stencil) { + if (this->has_stencil_attachment()) { + MTLAttachment stencil_attachment_prev = this->get_stencil_attachment(); + this->remove_stencil_attachment(); + this->add_stencil_attachment( + static_cast<gpu::MTLTexture *>(unwrap(attach.tex)), attach.mip, attach.layer); + this->set_stencil_attachment_clear_value( + stencil_attachment_prev.clear_value.stencil); + this->set_stencil_loadstore_op(stencil_attachment_prev.load_action, + stencil_attachment_prev.store_action); + } + else { + this->add_stencil_attachment( + static_cast<gpu::MTLTexture *>(unwrap(attach.tex)), attach.mip, attach.layer); + } + } + + /* Flag depth as added -- mirrors the behaviour in gl_framebuffer.cc to exit the for-loop + * after GPU_FB_DEPTH_STENCIL_ATTACHMENT has executed. */ + depth_added = true; + + if (first_attachment == GPU_FB_MAX_ATTACHMENT) { + /* Only use depth texture to get information if there is no color attachment. */ + first_attachment = type; + first_attachment_mtl = this->get_depth_attachment(); + } + } + else { + this->remove_depth_attachment(); + if (type == GPU_FB_DEPTH_STENCIL_ATTACHMENT && this->has_stencil_attachment()) { + this->remove_stencil_attachment(); + } + } + } break; + case GPU_FB_COLOR_ATTACHMENT0: + case GPU_FB_COLOR_ATTACHMENT1: + case GPU_FB_COLOR_ATTACHMENT2: + case GPU_FB_COLOR_ATTACHMENT3: + case GPU_FB_COLOR_ATTACHMENT4: + case GPU_FB_COLOR_ATTACHMENT5: { + int color_slot_ind = type - GPU_FB_COLOR_ATTACHMENT0; + if (attach.tex) { + /* If we already had a colour attachment, preserve load/clear-state parameters, + * but remove existing and add new attachment. */ + if (this->has_attachment_at_slot(color_slot_ind)) { + MTLAttachment color_attachment_prev = this->get_color_attachment(color_slot_ind); + + this->remove_color_attachment(color_slot_ind); + this->add_color_attachment(static_cast<gpu::MTLTexture *>(unwrap(attach.tex)), + color_slot_ind, + attach.mip, + attach.layer); + this->set_color_attachment_clear_color(color_slot_ind, + color_attachment_prev.clear_value.color); + this->set_color_loadstore_op(color_slot_ind, + color_attachment_prev.load_action, + color_attachment_prev.store_action); + } + else { + this->add_color_attachment(static_cast<gpu::MTLTexture *>(unwrap(attach.tex)), + color_slot_ind, + attach.mip, + attach.layer); + } + first_attachment = type; + first_attachment_mtl = this->get_color_attachment(color_slot_ind); + } + else { + this->remove_color_attachment(color_slot_ind); + } + } break; + default: + /* Non-attachment parameters. */ + break; + } + } + + /* Check whether the first attachment is SRGB. */ + if (first_attachment != GPU_FB_MAX_ATTACHMENT) { + is_srgb_ = (first_attachment_mtl.texture->format_get() == GPU_SRGB8_A8); + } + + /* Reset viewport and Scissor (If viewport is smaller or equal to the framebuffer size). */ + if (update_viewport && t_viewport[2] <= width_ && t_viewport[3] <= height_) { + + this->viewport_set(t_viewport); + this->scissor_set(t_viewport); + } + else { + this->viewport_reset(); + this->scissor_reset(); + } + + /* We have now updated our internal structures. */ + dirty_attachments_ = false; +} + +void MTLFrameBuffer::apply_state(void) +{ + MTLContext *mtl_ctx = static_cast<MTLContext *>(unwrap(GPU_context_active_get())); + BLI_assert(mtl_ctx); + if (mtl_ctx->active_fb == this) { + if (dirty_state_ == false && dirty_state_ctx_ == mtl_ctx) { + return; + } + + /* Ensure viewport has been set. Note: This should no longer happen, but kept for safety to + * track bugs. */ + if (viewport_[2] == 0 || viewport_[3] == 0) { + MTL_LOG_WARNING( + "Viewport had width and height of (0,0) -- Updating -- DEBUG Safety check\n"); + viewport_reset(); + } + + /* Update Context State. */ + mtl_ctx->set_viewport(viewport_[0], viewport_[1], viewport_[2], viewport_[3]); + mtl_ctx->set_scissor(scissor_[0], scissor_[1], scissor_[2], scissor_[3]); + mtl_ctx->set_scissor_enabled(scissor_test_); + + dirty_state_ = false; + dirty_state_ctx_ = mtl_ctx; + } + else { + MTL_LOG_ERROR( + "Attempting to set FrameBuffer State (VIEWPORT, SCISSOR), But FrameBuffer is not bound to " + "current Context.\n"); + } +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \ Adding and Removing attachments + * \{ */ + +bool MTLFrameBuffer::add_color_attachment(gpu::MTLTexture *texture, + uint slot, + int miplevel, + int layer) +{ + BLI_assert(this); + BLI_assert(slot >= 0 && slot < this->get_attachment_limit()); + + if (texture) { + if (miplevel < 0 || miplevel >= MTL_MAX_MIPMAP_COUNT) { + MTL_LOG_WARNING("Attachment specified with invalid mip level %u\n", miplevel); + miplevel = 0; + } + + /* Check if slot is in-use. */ + /* Assume attachment load by default. */ + colour_attachment_count_ += (!mtl_color_attachments_[slot].used) ? 1 : 0; + mtl_color_attachments_[slot].used = true; + mtl_color_attachments_[slot].texture = texture; + mtl_color_attachments_[slot].mip = miplevel; + mtl_color_attachments_[slot].load_action = GPU_LOADACTION_LOAD; + mtl_color_attachments_[slot].store_action = GPU_STOREACTION_STORE; + mtl_color_attachments_[slot].render_target_array_length = 0; + + /* Determine whether array slice or depth plane based on texture type. */ + switch (texture->type_) { + case GPU_TEXTURE_1D: + case GPU_TEXTURE_2D: + BLI_assert(layer <= 0); + mtl_color_attachments_[slot].slice = 0; + mtl_color_attachments_[slot].depth_plane = 0; + break; + case GPU_TEXTURE_1D_ARRAY: + if (layer < 0) { + layer = 0; + MTL_LOG_WARNING("TODO: Support layered rendering for 1D array textures, if needed.\n"); + } + BLI_assert(layer < texture->h_); + mtl_color_attachments_[slot].slice = layer; + mtl_color_attachments_[slot].depth_plane = 0; + break; + case GPU_TEXTURE_2D_ARRAY: + BLI_assert(layer < texture->d_); + mtl_color_attachments_[slot].slice = layer; + mtl_color_attachments_[slot].depth_plane = 0; + if (layer == -1) { + mtl_color_attachments_[slot].slice = 0; + mtl_color_attachments_[slot].render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_3D: + BLI_assert(layer < texture->d_); + mtl_color_attachments_[slot].slice = 0; + mtl_color_attachments_[slot].depth_plane = layer; + if (layer == -1) { + mtl_color_attachments_[slot].depth_plane = 0; + mtl_color_attachments_[slot].render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_CUBE: + BLI_assert(layer < 6); + mtl_color_attachments_[slot].slice = layer; + mtl_color_attachments_[slot].depth_plane = 0; + if (layer == -1) { + mtl_color_attachments_[slot].slice = 0; + mtl_color_attachments_[slot].depth_plane = 0; + mtl_color_attachments_[slot].render_target_array_length = 6; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_CUBE_ARRAY: + BLI_assert(layer < 6 * texture->d_); + /* TODO(Metal): Verify multilayered rendering for Cube arrays. */ + mtl_color_attachments_[slot].slice = layer; + mtl_color_attachments_[slot].depth_plane = 0; + if (layer == -1) { + mtl_color_attachments_[slot].slice = 0; + mtl_color_attachments_[slot].depth_plane = 0; + mtl_color_attachments_[slot].render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_BUFFER: + mtl_color_attachments_[slot].slice = 0; + mtl_color_attachments_[slot].depth_plane = 0; + break; + default: + MTL_LOG_ERROR("MTLFrameBuffer::add_color_attachment Unrecognised texture type %u\n", + texture->type_); + break; + } + + /* Update Framebuffer Resolution. */ + int width_of_miplayer, height_of_miplayer; + if (miplevel <= 0) { + width_of_miplayer = texture->width_get(); + height_of_miplayer = texture->height_get(); + } + else { + width_of_miplayer = max_ii(texture->width_get() >> miplevel, 1); + height_of_miplayer = max_ii(texture->height_get() >> miplevel, 1); + } + + if (width_ == 0 || height_ == 0) { + this->size_set(width_of_miplayer, height_of_miplayer); + this->scissor_reset(); + this->viewport_reset(); + BLI_assert(width_ > 0); + BLI_assert(height_ > 0); + } + else { + BLI_assert(width_ == width_of_miplayer); + BLI_assert(height_ == height_of_miplayer); + } + + /* Flag as dirty. */ + this->mark_dirty(); + } + else { + MTL_LOG_ERROR( + "Passing in null texture to MTLFrameBuffer::addColourAttachment (This could be due to not " + "all texture types being supported).\n"); + } + return true; +} + +bool MTLFrameBuffer::add_depth_attachment(gpu::MTLTexture *texture, int miplevel, int layer) +{ + BLI_assert(this); + + if (texture) { + if (miplevel < 0 || miplevel >= MTL_MAX_MIPMAP_COUNT) { + MTL_LOG_WARNING("Attachment specified with invalid mip level %u\n", miplevel); + miplevel = 0; + } + + /* Assume attachment load by default. */ + mtl_depth_attachment_.used = true; + mtl_depth_attachment_.texture = texture; + mtl_depth_attachment_.mip = miplevel; + mtl_depth_attachment_.load_action = GPU_LOADACTION_LOAD; + mtl_depth_attachment_.store_action = GPU_STOREACTION_STORE; + mtl_depth_attachment_.render_target_array_length = 0; + + /* Determine whether array slice or depth plane based on texture type. */ + switch (texture->type_) { + case GPU_TEXTURE_1D: + case GPU_TEXTURE_2D: + BLI_assert(layer <= 0); + mtl_depth_attachment_.slice = 0; + mtl_depth_attachment_.depth_plane = 0; + break; + case GPU_TEXTURE_1D_ARRAY: + if (layer < 0) { + layer = 0; + MTL_LOG_WARNING("TODO: Support layered rendering for 1D array textures, if needed\n"); + } + BLI_assert(layer < texture->h_); + mtl_depth_attachment_.slice = layer; + mtl_depth_attachment_.depth_plane = 0; + break; + case GPU_TEXTURE_2D_ARRAY: + BLI_assert(layer < texture->d_); + mtl_depth_attachment_.slice = layer; + mtl_depth_attachment_.depth_plane = 0; + if (layer == -1) { + mtl_depth_attachment_.slice = 0; + mtl_depth_attachment_.render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_3D: + BLI_assert(layer < texture->d_); + mtl_depth_attachment_.slice = 0; + mtl_depth_attachment_.depth_plane = layer; + if (layer == -1) { + mtl_depth_attachment_.depth_plane = 0; + mtl_depth_attachment_.render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_CUBE: + BLI_assert(layer < 6); + mtl_depth_attachment_.slice = layer; + mtl_depth_attachment_.depth_plane = 0; + if (layer == -1) { + mtl_depth_attachment_.slice = 0; + mtl_depth_attachment_.depth_plane = 0; + mtl_depth_attachment_.render_target_array_length = 1; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_CUBE_ARRAY: + /* TODO(Metal): Verify multilayered rendering for Cube arrays. */ + BLI_assert(layer < 6 * texture->d_); + mtl_depth_attachment_.slice = layer; + mtl_depth_attachment_.depth_plane = 0; + if (layer == -1) { + mtl_depth_attachment_.slice = 0; + mtl_depth_attachment_.depth_plane = 0; + mtl_depth_attachment_.render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_BUFFER: + mtl_depth_attachment_.slice = 0; + mtl_depth_attachment_.depth_plane = 0; + break; + default: + BLI_assert(false && "Unrecognised texture type"); + break; + } + + /* Update Framebuffer Resolution. */ + int width_of_miplayer, height_of_miplayer; + if (miplevel <= 0) { + width_of_miplayer = texture->width_get(); + height_of_miplayer = texture->height_get(); + } + else { + width_of_miplayer = max_ii(texture->width_get() >> miplevel, 1); + height_of_miplayer = max_ii(texture->height_get() >> miplevel, 1); + } + + /* Update Framebuffer Resolution. */ + if (width_ == 0 || height_ == 0) { + this->size_set(width_of_miplayer, height_of_miplayer); + this->scissor_reset(); + this->viewport_reset(); + BLI_assert(width_ > 0); + BLI_assert(height_ > 0); + } + else { + BLI_assert(width_ == texture->width_get()); + BLI_assert(height_ == texture->height_get()); + } + + /* Flag as dirty after attachments changed. */ + this->mark_dirty(); + } + else { + MTL_LOG_ERROR( + "Passing in null texture to MTLFrameBuffer::addDepthAttachment (This could be due to not " + "all texture types being supported)."); + } + return true; +} + +bool MTLFrameBuffer::add_stencil_attachment(gpu::MTLTexture *texture, int miplevel, int layer) +{ + BLI_assert(this); + + if (texture) { + if (miplevel < 0 || miplevel >= MTL_MAX_MIPMAP_COUNT) { + MTL_LOG_WARNING("Attachment specified with invalid mip level %u\n", miplevel); + miplevel = 0; + } + + /* Assume attachment load by default. */ + mtl_stencil_attachment_.used = true; + mtl_stencil_attachment_.texture = texture; + mtl_stencil_attachment_.mip = miplevel; + mtl_stencil_attachment_.load_action = GPU_LOADACTION_LOAD; + mtl_stencil_attachment_.store_action = GPU_STOREACTION_STORE; + mtl_stencil_attachment_.render_target_array_length = 0; + + /* Determine whether array slice or depth plane based on texture type. */ + switch (texture->type_) { + case GPU_TEXTURE_1D: + case GPU_TEXTURE_2D: + BLI_assert(layer <= 0); + mtl_stencil_attachment_.slice = 0; + mtl_stencil_attachment_.depth_plane = 0; + break; + case GPU_TEXTURE_1D_ARRAY: + if (layer < 0) { + layer = 0; + MTL_LOG_WARNING("TODO: Support layered rendering for 1D array textures, if needed\n"); + } + BLI_assert(layer < texture->h_); + mtl_stencil_attachment_.slice = layer; + mtl_stencil_attachment_.depth_plane = 0; + break; + case GPU_TEXTURE_2D_ARRAY: + BLI_assert(layer < texture->d_); + mtl_stencil_attachment_.slice = layer; + mtl_stencil_attachment_.depth_plane = 0; + if (layer == -1) { + mtl_stencil_attachment_.slice = 0; + mtl_stencil_attachment_.render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_3D: + BLI_assert(layer < texture->d_); + mtl_stencil_attachment_.slice = 0; + mtl_stencil_attachment_.depth_plane = layer; + if (layer == -1) { + mtl_stencil_attachment_.depth_plane = 0; + mtl_stencil_attachment_.render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_CUBE: + BLI_assert(layer < 6); + mtl_stencil_attachment_.slice = layer; + mtl_stencil_attachment_.depth_plane = 0; + if (layer == -1) { + mtl_stencil_attachment_.slice = 0; + mtl_stencil_attachment_.depth_plane = 0; + mtl_stencil_attachment_.render_target_array_length = 1; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_CUBE_ARRAY: + /* TODO(Metal): Verify multilayered rendering for Cube arrays. */ + BLI_assert(layer < 6 * texture->d_); + mtl_stencil_attachment_.slice = layer; + mtl_stencil_attachment_.depth_plane = 0; + if (layer == -1) { + mtl_stencil_attachment_.slice = 0; + mtl_stencil_attachment_.depth_plane = 0; + mtl_stencil_attachment_.render_target_array_length = texture->d_; + use_multilayered_rendering_ = true; + } + break; + case GPU_TEXTURE_BUFFER: + mtl_stencil_attachment_.slice = 0; + mtl_stencil_attachment_.depth_plane = 0; + break; + default: + BLI_assert(false && "Unrecognised texture type"); + break; + } + + /* Update Framebuffer Resolution. */ + int width_of_miplayer, height_of_miplayer; + if (miplevel <= 0) { + width_of_miplayer = texture->width_get(); + height_of_miplayer = texture->height_get(); + } + else { + width_of_miplayer = max_ii(texture->width_get() >> miplevel, 1); + height_of_miplayer = max_ii(texture->height_get() >> miplevel, 1); + } + + /* Update Framebuffer Resolution. */ + if (width_ == 0 || height_ == 0) { + this->size_set(width_of_miplayer, height_of_miplayer); + this->scissor_reset(); + this->viewport_reset(); + BLI_assert(width_ > 0); + BLI_assert(height_ > 0); + } + else { + BLI_assert(width_ == texture->width_get()); + BLI_assert(height_ == texture->height_get()); + } + + /* Flag as dirty after attachments changed. */ + this->mark_dirty(); + } + else { + MTL_LOG_ERROR( + "Passing in null texture to MTLFrameBuffer::addStencilAttachment (This could be due to " + "not all texture types being supported)."); + } + return true; +} + +bool MTLFrameBuffer::remove_color_attachment(uint slot) +{ + BLI_assert(this); + BLI_assert(slot >= 0 && slot < this->get_attachment_limit()); + + if (this->has_attachment_at_slot(slot)) { + colour_attachment_count_ -= (mtl_color_attachments_[slot].used) ? 1 : 0; + mtl_color_attachments_[slot].used = false; + this->ensure_render_target_size(); + this->mark_dirty(); + return true; + } + + return false; +} + +bool MTLFrameBuffer::remove_depth_attachment() +{ + BLI_assert(this); + + mtl_depth_attachment_.used = false; + mtl_depth_attachment_.texture = nullptr; + this->ensure_render_target_size(); + this->mark_dirty(); + + return true; +} + +bool MTLFrameBuffer::remove_stencil_attachment() +{ + BLI_assert(this); + + mtl_stencil_attachment_.used = false; + mtl_stencil_attachment_.texture = nullptr; + this->ensure_render_target_size(); + this->mark_dirty(); + + return true; +} + +void MTLFrameBuffer::remove_all_attachments() +{ + BLI_assert(this); + + for (int attachment = 0; attachment < GPU_FB_MAX_COLOR_ATTACHMENT; attachment++) { + this->remove_color_attachment(attachment); + } + this->remove_depth_attachment(); + this->remove_stencil_attachment(); + colour_attachment_count_ = 0; + this->mark_dirty(); + + /* Verify height. */ + this->ensure_render_target_size(); + + /* Flag attachments as no longer being dirty. */ + dirty_attachments_ = false; +} + +void MTLFrameBuffer::ensure_render_target_size() +{ + /* If we have no attachments, reset width and height to zero. */ + if (colour_attachment_count_ == 0 && !this->has_depth_attachment() && + !this->has_stencil_attachment()) { + + /* Reset Viewport and Scissor for NULL framebuffer. */ + this->size_set(0, 0); + this->scissor_reset(); + this->viewport_reset(); + } +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \ Clear values and Load-store actions + * \{ */ + +void MTLFrameBuffer::attachment_set_loadstore_op(GPUAttachmentType type, + eGPULoadOp load_action, + eGPUStoreOp store_action) +{ + if (type >= GPU_FB_COLOR_ATTACHMENT0) { + int slot = type - GPU_FB_COLOR_ATTACHMENT0; + this->set_color_loadstore_op(slot, load_action, store_action); + } + else if (type == GPU_FB_DEPTH_STENCIL_ATTACHMENT) { + this->set_depth_loadstore_op(load_action, store_action); + this->set_stencil_loadstore_op(load_action, store_action); + } + else if (type == GPU_FB_DEPTH_ATTACHMENT) { + this->set_depth_loadstore_op(load_action, store_action); + } +} + +bool MTLFrameBuffer::set_color_attachment_clear_color(uint slot, const float clear_color[4]) +{ + BLI_assert(this); + BLI_assert(slot >= 0 && slot < this->get_attachment_limit()); + + /* Only mark as dirty if values have changed. */ + bool changed = mtl_color_attachments_[slot].load_action != GPU_LOADACTION_CLEAR; + changed = changed || (memcmp(mtl_color_attachments_[slot].clear_value.color, + clear_color, + sizeof(float) * 4) != 0); + if (changed) { + memcpy(mtl_color_attachments_[slot].clear_value.color, clear_color, sizeof(float) * 4); + } + mtl_color_attachments_[slot].load_action = GPU_LOADACTION_CLEAR; + + if (changed) { + this->mark_loadstore_dirty(); + } + return true; +} + +bool MTLFrameBuffer::set_depth_attachment_clear_value(float depth_clear) +{ + BLI_assert(this); + + if (mtl_depth_attachment_.clear_value.depth != depth_clear || + mtl_depth_attachment_.load_action != GPU_LOADACTION_CLEAR) { + mtl_depth_attachment_.clear_value.depth = depth_clear; + mtl_depth_attachment_.load_action = GPU_LOADACTION_CLEAR; + this->mark_loadstore_dirty(); + } + return true; +} + +bool MTLFrameBuffer::set_stencil_attachment_clear_value(uint stencil_clear) +{ + BLI_assert(this); + + if (mtl_stencil_attachment_.clear_value.stencil != stencil_clear || + mtl_stencil_attachment_.load_action != GPU_LOADACTION_CLEAR) { + mtl_stencil_attachment_.clear_value.stencil = stencil_clear; + mtl_stencil_attachment_.load_action = GPU_LOADACTION_CLEAR; + this->mark_loadstore_dirty(); + } + return true; +} + +bool MTLFrameBuffer::set_color_loadstore_op(uint slot, + eGPULoadOp load_action, + eGPUStoreOp store_action) +{ + BLI_assert(this); + eGPULoadOp prev_load_action = mtl_color_attachments_[slot].load_action; + eGPUStoreOp prev_store_action = mtl_color_attachments_[slot].store_action; + mtl_color_attachments_[slot].load_action = load_action; + mtl_color_attachments_[slot].store_action = store_action; + + bool changed = (mtl_color_attachments_[slot].load_action != prev_load_action || + mtl_color_attachments_[slot].store_action != prev_store_action); + if (changed) { + this->mark_loadstore_dirty(); + } + + return changed; +} + +bool MTLFrameBuffer::set_depth_loadstore_op(eGPULoadOp load_action, eGPUStoreOp store_action) +{ + BLI_assert(this); + eGPULoadOp prev_load_action = mtl_depth_attachment_.load_action; + eGPUStoreOp prev_store_action = mtl_depth_attachment_.store_action; + mtl_depth_attachment_.load_action = load_action; + mtl_depth_attachment_.store_action = store_action; + + bool changed = (mtl_depth_attachment_.load_action != prev_load_action || + mtl_depth_attachment_.store_action != prev_store_action); + if (changed) { + this->mark_loadstore_dirty(); + } + + return changed; +} + +bool MTLFrameBuffer::set_stencil_loadstore_op(eGPULoadOp load_action, eGPUStoreOp store_action) +{ + BLI_assert(this); + eGPULoadOp prev_load_action = mtl_stencil_attachment_.load_action; + eGPUStoreOp prev_store_action = mtl_stencil_attachment_.store_action; + mtl_stencil_attachment_.load_action = load_action; + mtl_stencil_attachment_.store_action = store_action; + + bool changed = (mtl_stencil_attachment_.load_action != prev_load_action || + mtl_stencil_attachment_.store_action != prev_store_action); + if (changed) { + this->mark_loadstore_dirty(); + } + + return changed; +} + +bool MTLFrameBuffer::reset_clear_state() +{ + for (int slot = 0; slot < colour_attachment_count_; slot++) { + this->set_color_loadstore_op(slot, GPU_LOADACTION_LOAD, GPU_STOREACTION_STORE); + } + this->set_depth_loadstore_op(GPU_LOADACTION_LOAD, GPU_STOREACTION_STORE); + this->set_stencil_loadstore_op(GPU_LOADACTION_LOAD, GPU_STOREACTION_STORE); + return true; +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \ Fetch values and Framebuffer status + * \{ */ + +bool MTLFrameBuffer::has_attachment_at_slot(uint slot) +{ + BLI_assert(this); + + if (slot >= 0 && slot < this->get_attachment_limit()) { + return mtl_color_attachments_[slot].used; + } + return false; +} + +bool MTLFrameBuffer::has_color_attachment_with_texture(gpu::MTLTexture *texture) +{ + BLI_assert(this); + + for (int attachment = 0; attachment < this->get_attachment_limit(); attachment++) { + if (mtl_color_attachments_[attachment].used && + mtl_color_attachments_[attachment].texture == texture) { + return true; + } + } + return false; +} + +bool MTLFrameBuffer::has_depth_attachment() +{ + BLI_assert(this); + return mtl_depth_attachment_.used; +} + +bool MTLFrameBuffer::has_stencil_attachment() +{ + BLI_assert(this); + return mtl_stencil_attachment_.used; +} + +int MTLFrameBuffer::get_color_attachment_slot_from_texture(gpu::MTLTexture *texture) +{ + BLI_assert(this); + BLI_assert(texture); + + for (int attachment = 0; attachment < this->get_attachment_limit(); attachment++) { + if (mtl_color_attachments_[attachment].used && + (mtl_color_attachments_[attachment].texture == texture)) { + return attachment; + } + } + return -1; +} + +uint MTLFrameBuffer::get_attachment_count() +{ + BLI_assert(this); + return colour_attachment_count_; +} + +MTLAttachment MTLFrameBuffer::get_color_attachment(uint slot) +{ + BLI_assert(this); + if (slot >= 0 && slot < GPU_FB_MAX_COLOR_ATTACHMENT) { + return mtl_color_attachments_[slot]; + } + MTLAttachment null_attachment; + null_attachment.used = false; + return null_attachment; +} + +MTLAttachment MTLFrameBuffer::get_depth_attachment() +{ + BLI_assert(this); + return mtl_depth_attachment_; +} + +MTLAttachment MTLFrameBuffer::get_stencil_attachment() +{ + BLI_assert(this); + return mtl_stencil_attachment_; +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \ METAL API Resources and Validation + * \{ */ +bool MTLFrameBuffer::validate_render_pass() +{ + BLI_assert(this); + + /* First update attachments if dirty. */ + this->update_attachments(true); + + /* Verify attachment count. */ + int used_attachments = 0; + for (int attachment = 0; attachment < GPU_FB_MAX_COLOR_ATTACHMENT; attachment++) { + if (mtl_color_attachments_[attachment].used) { + used_attachments++; + } + } + used_attachments += (mtl_depth_attachment_.used) ? 1 : 0; + used_attachments += (mtl_stencil_attachment_.used) ? 1 : 0; + return (used_attachments > 0); +} + +MTLLoadAction mtl_load_action_from_gpu(eGPULoadOp action) +{ + return (action == GPU_LOADACTION_LOAD) ? + MTLLoadActionLoad : + ((action == GPU_LOADACTION_CLEAR) ? MTLLoadActionClear : MTLLoadActionDontCare); +} + +MTLStoreAction mtl_store_action_from_gpu(eGPUStoreOp action) +{ + return (action == GPU_STOREACTION_STORE) ? MTLStoreActionStore : MTLStoreActionDontCare; +} + +MTLRenderPassDescriptor *MTLFrameBuffer::bake_render_pass_descriptor(bool load_contents) +{ + BLI_assert(this); + if (load_contents) { + /* Only force-load contents if there is no clear pending. */ + BLI_assert(!has_pending_clear_); + } + + /* Ensure we are inside a frame boundary. */ + MTLContext *metal_ctx = static_cast<MTLContext *>(unwrap(GPU_context_active_get())); + BLI_assert(metal_ctx && metal_ctx->get_inside_frame()); + UNUSED_VARS_NDEBUG(metal_ctx); + + /* If Framebuffer has been modified, regenerate descriptor. */ + if (is_dirty_) { + /* Clear all configs. */ + for (int config = 0; config < 3; config++) { + descriptor_dirty_[config] = true; + } + } + else if (is_loadstore_dirty_) { + /* Load config always has load ops, so we only need to re-generate custom and clear state. */ + descriptor_dirty_[MTL_FB_CONFIG_CLEAR] = true; + descriptor_dirty_[MTL_FB_CONFIG_CUSTOM] = true; + } + + /* If we need to populate descriptor" */ + /* Select config based on FrameBuffer state: + * [0] {MTL_FB_CONFIG_CLEAR} = Clear config -- we have a pending clear so should perform our + * configured clear. + * [1] {MTL_FB_CONFIG_LOAD} = Load config -- We need to re-load ALL attachments, + * used for re-binding/pass-breaks. + * [2] {MTL_FB_CONFIG_CUSTOM} = Custom config -- Use this when a custom binding config is + * specified. + */ + uint descriptor_config = (load_contents) ? MTL_FB_CONFIG_LOAD : + ((this->get_pending_clear()) ? MTL_FB_CONFIG_CLEAR : + MTL_FB_CONFIG_CUSTOM); + if (descriptor_dirty_[descriptor_config] || framebuffer_descriptor_[descriptor_config] == nil) { + + /* Create descriptor if it does not exist. */ + if (framebuffer_descriptor_[descriptor_config] == nil) { + framebuffer_descriptor_[descriptor_config] = [[MTLRenderPassDescriptor alloc] init]; + } + +#if defined(MAC_OS_X_VERSION_11_0) && __MAC_OS_X_VERSION_MAX_ALLOWED > MAC_OS_X_VERSION_11_0 + if (@available(macOS 11.00, *)) { + /* Optimisation: Use smaller tile size on Apple Silicon if exceeding a certain bpp limit. */ + bool is_tile_based_gpu = [metal_ctx->device hasUnifiedMemory]; + if (is_tile_based_gpu) { + uint framebuffer_bpp = this->get_bits_per_pixel(); + bool use_small_tiles = (framebuffer_bpp > 64); + + if (use_small_tiles) { + framebuffer_descriptor_[descriptor_config].tileWidth = 16; + framebuffer_descriptor_[descriptor_config].tileHeight = 16; + } + } + } +#endif + + /* Configure multilayered rendering. */ + if (use_multilayered_rendering_) { + /* Ensure all targets have the same length. */ + int len = 0; + bool valid = true; + + for (int attachment_ind = 0; attachment_ind < GPU_FB_MAX_COLOR_ATTACHMENT; + attachment_ind++) { + if (mtl_color_attachments_[attachment_ind].used) { + if (len == 0) { + len = mtl_color_attachments_[attachment_ind].render_target_array_length; + } + else { + valid = valid && + (len == mtl_color_attachments_[attachment_ind].render_target_array_length); + } + } + } + + if (mtl_depth_attachment_.used) { + if (len == 0) { + len = mtl_depth_attachment_.render_target_array_length; + } + else { + valid = valid && (len == mtl_depth_attachment_.render_target_array_length); + } + } + + if (mtl_stencil_attachment_.used) { + if (len == 0) { + len = mtl_stencil_attachment_.render_target_array_length; + } + else { + valid = valid && (len == mtl_stencil_attachment_.render_target_array_length); + } + } + + BLI_assert(len > 0); + BLI_assert(valid); + framebuffer_descriptor_[descriptor_config].renderTargetArrayLength = len; + } + else { + framebuffer_descriptor_[descriptor_config].renderTargetArrayLength = 0; + } + + /* Colour attachments. */ + int colour_attachments = 0; + for (int attachment_ind = 0; attachment_ind < GPU_FB_MAX_COLOR_ATTACHMENT; attachment_ind++) { + + if (mtl_color_attachments_[attachment_ind].used) { + + /* Create attachment descriptor. */ + MTLRenderPassColorAttachmentDescriptor *attachment = + colour_attachment_descriptors_[attachment_ind]; + BLI_assert(attachment != nil); + + id<MTLTexture> texture = + mtl_color_attachments_[attachment_ind].texture->get_metal_handle_base(); + if (texture == nil) { + MTL_LOG_ERROR("Attempting to assign invalid texture as attachment\n"); + } + + /* IF SRGB is enabled, but we are rendering with SRGB disabled, sample texture view. */ + /* TODO(Metal): Consider caching SRGB texture view. */ + id<MTLTexture> source_color_texture = texture; + if (this->get_is_srgb() && !this->get_srgb_enabled()) { + source_color_texture = [texture newTextureViewWithPixelFormat:MTLPixelFormatRGBA8Unorm]; + } + + /* Resolve appropriate load action -- IF force load, perform load. + * If clear but framebuffer has no pending clear, also load. */ + eGPULoadOp load_action = mtl_color_attachments_[attachment_ind].load_action; + if (descriptor_config == MTL_FB_CONFIG_LOAD) { + /* MTL_FB_CONFIG_LOAD must always load. */ + load_action = GPU_LOADACTION_LOAD; + } + else if (descriptor_config == MTL_FB_CONFIG_CUSTOM && + load_action == GPU_LOADACTION_CLEAR) { + /* Custom config should be LOAD or DONT_CARE only. */ + load_action = GPU_LOADACTION_LOAD; + } + attachment.texture = source_color_texture; + attachment.loadAction = mtl_load_action_from_gpu(load_action); + attachment.clearColor = + (load_action == GPU_LOADACTION_CLEAR) ? + MTLClearColorMake(mtl_color_attachments_[attachment_ind].clear_value.color[0], + mtl_color_attachments_[attachment_ind].clear_value.color[1], + mtl_color_attachments_[attachment_ind].clear_value.color[2], + mtl_color_attachments_[attachment_ind].clear_value.color[3]) : + MTLClearColorMake(0.0, 0.0, 0.0, 0.0); + attachment.storeAction = mtl_store_action_from_gpu( + mtl_color_attachments_[attachment_ind].store_action); + attachment.level = mtl_color_attachments_[attachment_ind].mip; + attachment.slice = mtl_color_attachments_[attachment_ind].slice; + attachment.depthPlane = mtl_color_attachments_[attachment_ind].depth_plane; + colour_attachments++; + + /* Copy attachment info back in. */ + [framebuffer_descriptor_[descriptor_config].colorAttachments setObject:attachment + atIndexedSubscript:attachment_ind]; + } + else { + /* Disable colour attachment. */ + [framebuffer_descriptor_[descriptor_config].colorAttachments setObject:nil + atIndexedSubscript:attachment_ind]; + } + } + BLI_assert(colour_attachments == colour_attachment_count_); + + /* Depth attachment. */ + if (mtl_depth_attachment_.used) { + framebuffer_descriptor_[descriptor_config].depthAttachment.texture = + (id<MTLTexture>)mtl_depth_attachment_.texture->get_metal_handle_base(); + + /* Resolve appropriate load action -- IF force load, perform load. + * If clear but framebuffer has no pending clear, also load. */ + eGPULoadOp load_action = mtl_depth_attachment_.load_action; + if (descriptor_config == MTL_FB_CONFIG_LOAD) { + /* MTL_FB_CONFIG_LOAD must always load. */ + load_action = GPU_LOADACTION_LOAD; + } + else if (descriptor_config == MTL_FB_CONFIG_CUSTOM && load_action == GPU_LOADACTION_CLEAR) { + /* Custom config should be LOAD or DONT_CARE only. */ + load_action = GPU_LOADACTION_LOAD; + } + framebuffer_descriptor_[descriptor_config].depthAttachment.loadAction = + mtl_load_action_from_gpu(load_action); + framebuffer_descriptor_[descriptor_config].depthAttachment.clearDepth = + (load_action == GPU_LOADACTION_CLEAR) ? mtl_depth_attachment_.clear_value.depth : 0; + framebuffer_descriptor_[descriptor_config].depthAttachment.storeAction = + mtl_store_action_from_gpu(mtl_depth_attachment_.store_action); + framebuffer_descriptor_[descriptor_config].depthAttachment.level = mtl_depth_attachment_.mip; + framebuffer_descriptor_[descriptor_config].depthAttachment.slice = + mtl_depth_attachment_.slice; + framebuffer_descriptor_[descriptor_config].depthAttachment.depthPlane = + mtl_depth_attachment_.depth_plane; + } + else { + framebuffer_descriptor_[descriptor_config].depthAttachment.texture = nil; + } + + /* Stencil attachment. */ + if (mtl_stencil_attachment_.used) { + framebuffer_descriptor_[descriptor_config].stencilAttachment.texture = + (id<MTLTexture>)mtl_stencil_attachment_.texture->get_metal_handle_base(); + + /* Resolve appropriate load action -- IF force load, perform load. + * If clear but framebuffer has no pending clear, also load. */ + eGPULoadOp load_action = mtl_stencil_attachment_.load_action; + if (descriptor_config == MTL_FB_CONFIG_LOAD) { + /* MTL_FB_CONFIG_LOAD must always load. */ + load_action = GPU_LOADACTION_LOAD; + } + else if (descriptor_config == MTL_FB_CONFIG_CUSTOM && load_action == GPU_LOADACTION_CLEAR) { + /* Custom config should be LOAD or DONT_CARE only. */ + load_action = GPU_LOADACTION_LOAD; + } + framebuffer_descriptor_[descriptor_config].stencilAttachment.loadAction = + mtl_load_action_from_gpu(load_action); + framebuffer_descriptor_[descriptor_config].stencilAttachment.clearStencil = + (load_action == GPU_LOADACTION_CLEAR) ? mtl_stencil_attachment_.clear_value.stencil : 0; + framebuffer_descriptor_[descriptor_config].stencilAttachment.storeAction = + mtl_store_action_from_gpu(mtl_stencil_attachment_.store_action); + framebuffer_descriptor_[descriptor_config].stencilAttachment.level = + mtl_stencil_attachment_.mip; + framebuffer_descriptor_[descriptor_config].stencilAttachment.slice = + mtl_stencil_attachment_.slice; + framebuffer_descriptor_[descriptor_config].stencilAttachment.depthPlane = + mtl_stencil_attachment_.depth_plane; + } + else { + framebuffer_descriptor_[descriptor_config].stencilAttachment.texture = nil; + } + descriptor_dirty_[descriptor_config] = false; + } + is_dirty_ = false; + is_loadstore_dirty_ = false; + return framebuffer_descriptor_[descriptor_config]; +} + +/** \} */ + +/* -------------------------------------------------------------------- */ +/** \ Blitting + * \{ */ + +void MTLFrameBuffer::blit(uint read_slot, + uint src_x_offset, + uint src_y_offset, + MTLFrameBuffer *metal_fb_write, + uint write_slot, + uint dst_x_offset, + uint dst_y_offset, + uint width, + uint height, + eGPUFrameBufferBits blit_buffers) +{ + BLI_assert(this); + BLI_assert(metal_fb_write); + if (!(this && metal_fb_write)) { + return; + } + MTLContext *mtl_context = reinterpret_cast<MTLContext *>(GPU_context_active_get()); + + const bool do_color = (blit_buffers & GPU_COLOR_BIT); + const bool do_depth = (blit_buffers & GPU_DEPTH_BIT); + const bool do_stencil = (blit_buffers & GPU_STENCIL_BIT); + + /* Early exit if there is no blit to do. */ + if (!(do_color || do_depth || do_stencil)) { + MTL_LOG_WARNING( + " MTLFrameBuffer: requested blit but no color, depth or stencil flag was set\n"); + return; + } + + id<MTLBlitCommandEncoder> blit_encoder = nil; + + /* If the color format is not the same, we cannot use the BlitCommandEncoder, and instead use + * a Graphics-based blit. */ + if (do_color && (this->get_color_attachment(read_slot).texture->format_get() != + metal_fb_write->get_color_attachment(read_slot).texture->format_get())) { + + MTLAttachment src_attachment = this->get_color_attachment(read_slot); + MTLAttachment dst_attachment = metal_fb_write->get_color_attachment(write_slot); + assert(src_attachment.slice == 0 && + "currently only supporting slice 0 for graphics framebuffer blit"); + + src_attachment.texture->blit(dst_attachment.texture, + src_x_offset, + src_y_offset, + dst_x_offset, + dst_y_offset, + src_attachment.mip, + dst_attachment.mip, + dst_attachment.slice, + width, + height); + } + else { + + /* Setup blit encoder. */ + blit_encoder = mtl_context->main_command_buffer.ensure_begin_blit_encoder(); + + if (do_color) { + MTLAttachment src_attachment = this->get_color_attachment(read_slot); + MTLAttachment dst_attachment = metal_fb_write->get_color_attachment(write_slot); + + if (src_attachment.used && dst_attachment.used) { + + /* TODO(Metal): Support depth(z) offset in blit if needed. */ + src_attachment.texture->blit(blit_encoder, + src_x_offset, + src_y_offset, + 0, + src_attachment.slice, + src_attachment.mip, + dst_attachment.texture, + dst_x_offset, + dst_y_offset, + 0, + dst_attachment.slice, + dst_attachment.mip, + width, + height, + 1); + } + else { + MTL_LOG_ERROR("Failed performing colour blit\n"); + } + } + } + if ((do_depth || do_stencil) && blit_encoder == nil) { + blit_encoder = mtl_context->main_command_buffer.ensure_begin_blit_encoder(); + } + + if (do_depth) { + MTLAttachment src_attachment = this->get_depth_attachment(); + MTLAttachment dst_attachment = metal_fb_write->get_depth_attachment(); + + if (src_attachment.used && dst_attachment.used) { + + /* TODO(Metal): Support depth(z) offset in blit if needed. */ + src_attachment.texture->blit(blit_encoder, + src_x_offset, + src_y_offset, + 0, + src_attachment.slice, + src_attachment.mip, + dst_attachment.texture, + dst_x_offset, + dst_y_offset, + 0, + dst_attachment.slice, + dst_attachment.mip, + width, + height, + 1); + } + else { + MTL_LOG_ERROR("Failed performing depth blit\n"); + } + } + + /* Stencil attachment blit. */ + if (do_stencil) { + MTLAttachment src_attachment = this->get_stencil_attachment(); + MTLAttachment dst_attachment = metal_fb_write->get_stencil_attachment(); + + if (src_attachment.used && dst_attachment.used) { + + /* TODO(Metal): Support depth(z) offset in blit if needed. */ + src_attachment.texture->blit(blit_encoder, + src_x_offset, + src_y_offset, + 0, + src_attachment.slice, + src_attachment.mip, + dst_attachment.texture, + dst_x_offset, + dst_y_offset, + 0, + dst_attachment.slice, + dst_attachment.mip, + width, + height, + 1); + } + else { + MTL_LOG_ERROR("Failed performing Stencil blit\n"); + } + } +} + +int MTLFrameBuffer::get_width() +{ + return width_; +} +int MTLFrameBuffer::get_height() +{ + return height_; +} + +} // blender::gpu diff --git a/source/blender/gpu/metal/mtl_state.hh b/source/blender/gpu/metal/mtl_state.hh index f2d85f9648b..23bf8600ddd 100644 --- a/source/blender/gpu/metal/mtl_state.hh +++ b/source/blender/gpu/metal/mtl_state.hh @@ -62,10 +62,10 @@ class MTLStateManager : public StateManager { void set_mutable_state(const GPUStateMutable &state); /* METAL State utility functions. */ - void mtl_state_init(void); + void mtl_state_init(); void mtl_depth_range(float near, float far); - void mtl_stencil_mask(unsigned int mask); - void mtl_stencil_set_func(eGPUStencilTest stencil_func, int ref, unsigned int mask); + void mtl_stencil_mask(uint mask); + void mtl_stencil_set_func(eGPUStencilTest stencil_func, int ref, uint mask); MEM_CXX_CLASS_ALLOC_FUNCS("MTLStateManager") }; diff --git a/source/blender/gpu/metal/mtl_state.mm b/source/blender/gpu/metal/mtl_state.mm index fa2f5c54391..cf7fbdba6b9 100644 --- a/source/blender/gpu/metal/mtl_state.mm +++ b/source/blender/gpu/metal/mtl_state.mm @@ -8,6 +8,7 @@ #include "GPU_framebuffer.h" #include "mtl_context.hh" +#include "mtl_framebuffer.hh" #include "mtl_state.hh" namespace blender::gpu { @@ -18,14 +19,14 @@ namespace blender::gpu { void MTLStateManager::mtl_state_init(void) { - BLI_assert(this->context_); - this->context_->pipeline_state_init(); + BLI_assert(context_); + context_->pipeline_state_init(); } MTLStateManager::MTLStateManager(MTLContext *ctx) : StateManager() { /* Initialize State. */ - this->context_ = ctx; + context_ = ctx; mtl_state_init(); /* Force update using default state. */ @@ -39,8 +40,9 @@ void MTLStateManager::apply_state(void) { this->set_state(this->state); this->set_mutable_state(this->mutable_state); - /* TODO(Metal): Enable after integration of MTLFrameBuffer. */ - /* static_cast<MTLFrameBuffer *>(this->context_->active_fb)->apply_state(); */ + + /* Apply active FrameBuffer state. */ + static_cast<MTLFrameBuffer *>(context_->active_fb)->apply_state(); }; void MTLStateManager::force_state(void) @@ -103,10 +105,10 @@ void MTLStateManager::set_state(const GPUState &state) void MTLStateManager::mtl_depth_range(float near, float far) { - BLI_assert(this->context_); + BLI_assert(context_); BLI_assert(near >= 0.0 && near < 1.0); BLI_assert(far > 0.0 && far <= 1.0); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; MTLContextDepthStencilState &ds_state = pipeline_state.depth_stencil_state; ds_state.depth_range_near = near; @@ -117,7 +119,7 @@ void MTLStateManager::mtl_depth_range(float near, float far) void MTLStateManager::set_mutable_state(const GPUStateMutable &state) { GPUStateMutable changed = state ^ current_mutable_; - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; if (float_as_uint(changed.point_size) != 0) { pipeline_state.point_size = state.point_size; @@ -150,8 +152,8 @@ void MTLStateManager::set_mutable_state(const GPUStateMutable &state) void MTLStateManager::set_write_mask(const eGPUWriteMask value) { - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; pipeline_state.depth_stencil_state.depth_write_enable = ((value & GPU_WRITE_DEPTH) != 0); pipeline_state.color_write_mask = (((value & GPU_WRITE_RED) != 0) ? MTLColorWriteMaskRed : MTLColorWriteMaskNone) | @@ -205,8 +207,8 @@ static MTLCompareFunction gpu_stencil_func_to_metal(eGPUStencilTest stencil_func void MTLStateManager::set_depth_test(const eGPUDepthTest value) { - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; MTLContextDepthStencilState &ds_state = pipeline_state.depth_stencil_state; ds_state.depth_test_enabled = (value != GPU_DEPTH_NONE); @@ -214,20 +216,18 @@ void MTLStateManager::set_depth_test(const eGPUDepthTest value) pipeline_state.dirty_flags |= MTL_PIPELINE_STATE_DEPTHSTENCIL_FLAG; } -void MTLStateManager::mtl_stencil_mask(unsigned int mask) +void MTLStateManager::mtl_stencil_mask(uint mask) { - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; pipeline_state.depth_stencil_state.stencil_write_mask = mask; pipeline_state.dirty_flags |= MTL_PIPELINE_STATE_DEPTHSTENCIL_FLAG; } -void MTLStateManager::mtl_stencil_set_func(eGPUStencilTest stencil_func, - int ref, - unsigned int mask) +void MTLStateManager::mtl_stencil_set_func(eGPUStencilTest stencil_func, int ref, uint mask) { - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; MTLContextDepthStencilState &ds_state = pipeline_state.depth_stencil_state; ds_state.stencil_func = gpu_stencil_func_to_metal(stencil_func); @@ -275,19 +275,17 @@ void MTLStateManager::set_stencil_test(const eGPUStencilTest test, const eGPUSte { switch (operation) { case GPU_STENCIL_OP_REPLACE: - mtl_stencil_set_op(this->context_, - MTLStencilOperationKeep, - MTLStencilOperationKeep, - MTLStencilOperationReplace); + mtl_stencil_set_op( + context_, MTLStencilOperationKeep, MTLStencilOperationKeep, MTLStencilOperationReplace); break; case GPU_STENCIL_OP_COUNT_DEPTH_PASS: /* Winding inversed due to flipped Y coordinate system in Metal. */ - mtl_stencil_set_op_separate(this->context_, + mtl_stencil_set_op_separate(context_, GPU_CULL_FRONT, MTLStencilOperationKeep, MTLStencilOperationKeep, MTLStencilOperationIncrementWrap); - mtl_stencil_set_op_separate(this->context_, + mtl_stencil_set_op_separate(context_, GPU_CULL_BACK, MTLStencilOperationKeep, MTLStencilOperationKeep, @@ -295,12 +293,12 @@ void MTLStateManager::set_stencil_test(const eGPUStencilTest test, const eGPUSte break; case GPU_STENCIL_OP_COUNT_DEPTH_FAIL: /* Winding inversed due to flipped Y coordinate system in Metal. */ - mtl_stencil_set_op_separate(this->context_, + mtl_stencil_set_op_separate(context_, GPU_CULL_FRONT, MTLStencilOperationKeep, MTLStencilOperationDecrementWrap, MTLStencilOperationKeep); - mtl_stencil_set_op_separate(this->context_, + mtl_stencil_set_op_separate(context_, GPU_CULL_BACK, MTLStencilOperationKeep, MTLStencilOperationIncrementWrap, @@ -308,14 +306,12 @@ void MTLStateManager::set_stencil_test(const eGPUStencilTest test, const eGPUSte break; case GPU_STENCIL_OP_NONE: default: - mtl_stencil_set_op(this->context_, - MTLStencilOperationKeep, - MTLStencilOperationKeep, - MTLStencilOperationKeep); + mtl_stencil_set_op( + context_, MTLStencilOperationKeep, MTLStencilOperationKeep, MTLStencilOperationKeep); } - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; pipeline_state.depth_stencil_state.stencil_test_enabled = (test != GPU_STENCIL_NONE); pipeline_state.dirty_flags |= MTL_PIPELINE_STATE_DEPTHSTENCIL_FLAG; } @@ -347,8 +343,8 @@ void MTLStateManager::set_logic_op(const bool enable) void MTLStateManager::set_facing(const bool invert) { /* Check Current Context. */ - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; /* Apply State -- opposite of GL, as METAL default is GPU_CLOCKWISE, GL default is * COUNTERCLOCKWISE. This needs to be the inverse of the default. */ @@ -362,8 +358,8 @@ void MTLStateManager::set_facing(const bool invert) void MTLStateManager::set_backface_culling(const eGPUFaceCullTest test) { /* Check Current Context. */ - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; /* Apply State. */ pipeline_state.culling_enabled = (test != GPU_CULL_NONE); @@ -386,8 +382,8 @@ void MTLStateManager::set_provoking_vert(const eGPUProvokingVertex vert) void MTLStateManager::set_shadow_bias(const bool enable) { /* Check Current Context. */ - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; MTLContextDepthStencilState &ds_state = pipeline_state.depth_stencil_state; /* Apply State. */ @@ -500,8 +496,8 @@ void MTLStateManager::set_blend(const eGPUBlend value) } /* Check Current Context. */ - BLI_assert(this->context_); - MTLContextGlobalShaderPipelineState &pipeline_state = this->context_->pipeline_state; + BLI_assert(context_); + MTLContextGlobalShaderPipelineState &pipeline_state = context_->pipeline_state; if (value == GPU_BLEND_SUBTRACT) { pipeline_state.rgb_blend_op = MTLBlendOperationReverseSubtract; @@ -549,58 +545,18 @@ void MTLStateManager::issue_barrier(eGPUBarrier barrier_bits) MTLContext *ctx = reinterpret_cast<MTLContext *>(GPU_context_active_get()); BLI_assert(ctx); - if (ctx->is_render_pass_active()) { - - /* Apple Silicon does not support memory barriers. - * We do not currently need these due to implicit API guarantees. - * NOTE(Metal): MTLFence/MTLEvent may be required to synchronize work if - * untracked resources are ever used. */ - if ([ctx->device hasUnifiedMemory]) { - return; - } - /* Issue barrier. */ - /* TODO(Metal): To be completed pending implementation of RenderCommandEncoder management. */ - id<MTLRenderCommandEncoder> rec = nil; // ctx->get_active_render_command_encoder(); - BLI_assert(rec); - - /* Only supporting Metal on 10.15 onward anyway - Check required for warnings. */ - if (@available(macOS 10.14, *)) { - MTLBarrierScope scope = 0; - if (barrier_bits & GPU_BARRIER_SHADER_IMAGE_ACCESS || - barrier_bits & GPU_BARRIER_TEXTURE_FETCH) { - scope = scope | MTLBarrierScopeTextures | MTLBarrierScopeRenderTargets; - } - if (barrier_bits & GPU_BARRIER_SHADER_STORAGE || - barrier_bits & GPU_BARRIER_VERTEX_ATTRIB_ARRAY || - barrier_bits & GPU_BARRIER_ELEMENT_ARRAY) { - scope = scope | MTLBarrierScopeBuffers; - } - - MTLRenderStages before_stage_flags = 0; - MTLRenderStages after_stage_flags = 0; - if (before_stages & GPU_BARRIER_STAGE_VERTEX && - !(before_stages & GPU_BARRIER_STAGE_FRAGMENT)) { - before_stage_flags = before_stage_flags | MTLRenderStageVertex; - } - if (before_stages & GPU_BARRIER_STAGE_FRAGMENT) { - before_stage_flags = before_stage_flags | MTLRenderStageFragment; - } - if (after_stages & GPU_BARRIER_STAGE_VERTEX) { - after_stage_flags = after_stage_flags | MTLRenderStageVertex; - } - if (after_stages & GPU_BARRIER_STAGE_FRAGMENT) { - after_stage_flags = MTLRenderStageFragment; - } - - if (scope != 0) { - [rec memoryBarrierWithScope:scope - afterStages:after_stage_flags - beforeStages:before_stage_flags]; - } - } + /* Apple Silicon does not support memory barriers. + * We do not currently need these due to implicit API guarantees. + * Note(Metal): MTLFence/MTLEvent may be required to synchronize work if + * untracked resources are ever used. */ + if ([ctx->device hasUnifiedMemory]) { + return; } + + ctx->main_command_buffer.insert_memory_barrier(barrier_bits, before_stages, after_stages); } + /** \} */ /* -------------------------------------------------------------------- */ diff --git a/source/blender/gpu/metal/mtl_texture.hh b/source/blender/gpu/metal/mtl_texture.hh index b4b1e91c496..0f908995a93 100644 --- a/source/blender/gpu/metal/mtl_texture.hh +++ b/source/blender/gpu/metal/mtl_texture.hh @@ -40,7 +40,7 @@ struct TextureUpdateRoutineSpecialisation { /* Number of channels the destination texture has (min=1, max=4). */ int component_count_output; - inline bool operator==(const TextureUpdateRoutineSpecialisation &other) const + bool operator==(const TextureUpdateRoutineSpecialisation &other) const { return ((input_data_type == other.input_data_type) && (output_data_type == other.output_data_type) && @@ -48,7 +48,7 @@ struct TextureUpdateRoutineSpecialisation { (component_count_output == other.component_count_output)); } - inline uint64_t hash() const + uint64_t hash() const { blender::DefaultHash<std::string> string_hasher; return (uint64_t)string_hasher( @@ -71,12 +71,12 @@ typedef enum { struct DepthTextureUpdateRoutineSpecialisation { DepthTextureUpdateMode data_mode; - inline bool operator==(const DepthTextureUpdateRoutineSpecialisation &other) const + bool operator==(const DepthTextureUpdateRoutineSpecialisation &other) const { return ((data_mode == other.data_mode)); } - inline uint64_t hash() const + uint64_t hash() const { return (uint64_t)(this->data_mode); } @@ -93,10 +93,10 @@ struct TextureReadRoutineSpecialisation { * 0 = Not a Depth format, * 1 = FLOAT DEPTH, * 2 = 24Bit Integer Depth, - * 4 = 32bit unsigned Integer Depth. */ + * 4 = 32bit uinteger Depth. */ int depth_format_mode; - inline bool operator==(const TextureReadRoutineSpecialisation &other) const + bool operator==(const TextureReadRoutineSpecialisation &other) const { return ((input_data_type == other.input_data_type) && (output_data_type == other.output_data_type) && @@ -105,7 +105,7 @@ struct TextureReadRoutineSpecialisation { (depth_format_mode == other.depth_format_mode)); } - inline uint64_t hash() const + uint64_t hash() const { blender::DefaultHash<std::string> string_hasher; return (uint64_t)string_hasher(this->input_data_type + this->output_data_type + @@ -125,28 +125,27 @@ static const int MTL_MAX_MIPMAP_COUNT = 15; /* Max: 16384x16384 */ static const int MTL_MAX_FBO_ATTACHED = 16; /* Samplers */ -typedef struct MTLSamplerState { +struct MTLSamplerState { eGPUSamplerState state; /* Mip min and mip max on sampler state always the same. * Level range now controlled with textureView to be consistent with GL baseLevel. */ - inline bool operator==(const MTLSamplerState &other) const + bool operator==(const MTLSamplerState &other) const { /* Add other parameters as needed. */ return (this->state == other.state); } - operator unsigned int() const + operator uint() const { - return (unsigned int)state; + return (uint)state; } operator uint64_t() const { return (uint64_t)state; } - -} MTLSamplerState; +}; const MTLSamplerState DEFAULT_SAMPLER_STATE = {GPU_SAMPLER_DEFAULT /*, 0, 9999*/}; @@ -174,12 +173,12 @@ class MTLTexture : public Texture { /* Texture Storage. */ id<MTLBuffer> texture_buffer_; - unsigned int aligned_w_ = 0; + uint aligned_w_ = 0; /* Blit Frame-buffer. */ GPUFrameBuffer *blit_fb_ = nullptr; - unsigned int blit_fb_slice_ = 0; - unsigned int blit_fb_mip_ = 0; + uint blit_fb_slice_ = 0; + uint blit_fb_mip_ = 0; /* Texture view properties */ /* In Metal, we use texture views to either limit mipmap ranges, @@ -252,7 +251,7 @@ class MTLTexture : public Texture { uint gl_bindcode_get(void) const override; bool texture_is_baked(); - inline const char *get_name() + const char *get_name() { return name_; } @@ -280,7 +279,7 @@ class MTLTexture : public Texture { void ensure_mipmaps(int miplvl); /* Flags a given mip level as being used. */ - void add_subresource(unsigned int level); + void add_subresource(uint level); void read_internal(int mip, int x_off, @@ -299,31 +298,31 @@ class MTLTexture : public Texture { id<MTLTexture> get_metal_handle_base(); MTLSamplerState get_sampler_state(); void blit(id<MTLBlitCommandEncoder> blit_encoder, - unsigned int src_x_offset, - unsigned int src_y_offset, - unsigned int src_z_offset, - unsigned int src_slice, - unsigned int src_mip, + uint src_x_offset, + uint src_y_offset, + uint src_z_offset, + uint src_slice, + uint src_mip, gpu::MTLTexture *dest, - unsigned int dst_x_offset, - unsigned int dst_y_offset, - unsigned int dst_z_offset, - unsigned int dst_slice, - unsigned int dst_mip, - unsigned int width, - unsigned int height, - unsigned int depth); + uint dst_x_offset, + uint dst_y_offset, + uint dst_z_offset, + uint dst_slice, + uint dst_mip, + uint width, + uint height, + uint depth); void blit(gpu::MTLTexture *dest, - unsigned int src_x_offset, - unsigned int src_y_offset, - unsigned int dst_x_offset, - unsigned int dst_y_offset, - unsigned int src_mip, - unsigned int dst_mip, - unsigned int dst_slice, + uint src_x_offset, + uint src_y_offset, + uint dst_x_offset, + uint dst_y_offset, + uint src_mip, + uint dst_mip, + uint dst_slice, int width, int height); - GPUFrameBuffer *get_blit_framebuffer(unsigned int dst_slice, unsigned int dst_mip); + GPUFrameBuffer *get_blit_framebuffer(uint dst_slice, uint dst_mip); MEM_CXX_CLASS_ALLOC_FUNCS("gpu::MTLTexture") diff --git a/source/blender/gpu/metal/mtl_texture.mm b/source/blender/gpu/metal/mtl_texture.mm index ca19d1f9e4b..ff2c2fce235 100644 --- a/source/blender/gpu/metal/mtl_texture.mm +++ b/source/blender/gpu/metal/mtl_texture.mm @@ -23,13 +23,6 @@ #include "GHOST_C-api.h" -/* Debug assistance. */ -/* Capture texture update routine for analysis in XCode GPU Frame Debugger. */ -#define DEBUG_TEXTURE_UPDATE_CAPTURE false - -/* Capture texture read routine for analysis in XCode GPU Frame Debugger. */ -#define DEBUG_TEXTURE_READ_CAPTURE false - namespace blender::gpu { /* -------------------------------------------------------------------- */ @@ -41,34 +34,34 @@ void gpu::MTLTexture::mtl_texture_init() BLI_assert(MTLContext::get() != nullptr); /* Status. */ - this->is_baked_ = false; - this->is_dirty_ = false; - this->resource_mode_ = MTL_TEXTURE_MODE_DEFAULT; - this->mtl_max_mips_ = 1; + is_baked_ = false; + is_dirty_ = false; + resource_mode_ = MTL_TEXTURE_MODE_DEFAULT; + mtl_max_mips_ = 1; /* Metal properties. */ - this->texture_ = nil; - this->texture_buffer_ = nil; - this->mip_swizzle_view_ = nil; + texture_ = nil; + texture_buffer_ = nil; + mip_swizzle_view_ = nil; /* Binding information. */ - this->is_bound_ = false; + is_bound_ = false; /* VBO. */ - this->vert_buffer_ = nullptr; - this->vert_buffer_mtl_ = nil; - this->vert_buffer_offset_ = -1; + vert_buffer_ = nullptr; + vert_buffer_mtl_ = nil; + vert_buffer_offset_ = -1; /* Default Swizzle. */ - this->tex_swizzle_mask_[0] = 'r'; - this->tex_swizzle_mask_[1] = 'g'; - this->tex_swizzle_mask_[2] = 'b'; - this->tex_swizzle_mask_[3] = 'a'; - this->mtl_swizzle_mask_ = MTLTextureSwizzleChannelsMake( + tex_swizzle_mask_[0] = 'r'; + tex_swizzle_mask_[1] = 'g'; + tex_swizzle_mask_[2] = 'b'; + tex_swizzle_mask_[3] = 'a'; + mtl_swizzle_mask_ = MTLTextureSwizzleChannelsMake( MTLTextureSwizzleRed, MTLTextureSwizzleGreen, MTLTextureSwizzleBlue, MTLTextureSwizzleAlpha); /* TODO(Metal): Find a way of specifying texture usage externally. */ - this->gpu_image_usage_flags_ = GPU_TEXTURE_USAGE_SHADER_READ | GPU_TEXTURE_USAGE_ATTACHMENT; + gpu_image_usage_flags_ = GPU_TEXTURE_USAGE_SHADER_READ | GPU_TEXTURE_USAGE_ATTACHMENT; } gpu::MTLTexture::MTLTexture(const char *name) : Texture(name) @@ -89,23 +82,23 @@ gpu::MTLTexture::MTLTexture(const char *name, /* Prep texture from METAL handle. */ BLI_assert(metal_texture != nil); BLI_assert(type == GPU_TEXTURE_2D); - this->type_ = type; + type_ = type; init_2D(metal_texture.width, metal_texture.height, 0, 1, format); /* Assign MTLTexture. */ - this->texture_ = metal_texture; - [this->texture_ retain]; + texture_ = metal_texture; + [texture_ retain]; /* Flag as Baked. */ - this->is_baked_ = true; - this->is_dirty_ = false; - this->resource_mode_ = MTL_TEXTURE_MODE_EXTERNAL; + is_baked_ = true; + is_dirty_ = false; + resource_mode_ = MTL_TEXTURE_MODE_EXTERNAL; } gpu::MTLTexture::~MTLTexture() { /* Unbind if bound. */ - if (this->is_bound_) { + if (is_bound_) { MTLContext *ctx = static_cast<MTLContext *>(unwrap(GPU_context_active_get())); if (ctx != nullptr) { ctx->state_manager->texture_unbind(this); @@ -123,49 +116,49 @@ void gpu::MTLTexture::bake_mip_swizzle_view() { if (texture_view_dirty_flags_) { /* if a texture view was previously created we release it. */ - if (this->mip_swizzle_view_ != nil) { - [this->mip_swizzle_view_ release]; + if (mip_swizzle_view_ != nil) { + [mip_swizzle_view_ release]; + mip_swizzle_view_ = nil; } /* Determine num slices */ int num_slices = 1; - switch (this->type_) { + switch (type_) { case GPU_TEXTURE_1D_ARRAY: - num_slices = this->h_; + num_slices = h_; break; case GPU_TEXTURE_2D_ARRAY: - num_slices = this->d_; + num_slices = d_; break; case GPU_TEXTURE_CUBE: num_slices = 6; break; case GPU_TEXTURE_CUBE_ARRAY: /* d_ is equal to array levels * 6, including face count. */ - num_slices = this->d_; + num_slices = d_; break; default: num_slices = 1; break; } - int range_len = min_ii((this->mip_texture_max_level_ - this->mip_texture_base_level_) + 1, - this->texture_.mipmapLevelCount); + int range_len = min_ii((mip_texture_max_level_ - mip_texture_base_level_) + 1, + texture_.mipmapLevelCount); BLI_assert(range_len > 0); - BLI_assert(mip_texture_base_level_ < this->texture_.mipmapLevelCount); - BLI_assert(this->mip_texture_base_layer_ < num_slices); - this->mip_swizzle_view_ = [this->texture_ - newTextureViewWithPixelFormat:this->texture_.pixelFormat - textureType:this->texture_.textureType - levels:NSMakeRange(this->mip_texture_base_level_, range_len) - slices:NSMakeRange(this->mip_texture_base_layer_, num_slices) - swizzle:this->mtl_swizzle_mask_]; + BLI_assert(mip_texture_base_level_ < texture_.mipmapLevelCount); + BLI_assert(mip_texture_base_layer_ < num_slices); + mip_swizzle_view_ = [texture_ + newTextureViewWithPixelFormat:texture_.pixelFormat + textureType:texture_.textureType + levels:NSMakeRange(mip_texture_base_level_, range_len) + slices:NSMakeRange(mip_texture_base_layer_, num_slices) + swizzle:mtl_swizzle_mask_]; MTL_LOG_INFO( "Updating texture view - MIP TEXTURE BASE LEVEL: %d, MAX LEVEL: %d (Range len: %d)\n", - this->mip_texture_base_level_, - min_ii(this->mip_texture_max_level_, this->texture_.mipmapLevelCount), + mip_texture_base_level_, + min_ii(mip_texture_max_level_, texture_.mipmapLevelCount), range_len); - [this->mip_swizzle_view_ retain]; - this->mip_swizzle_view_.label = [this->texture_ label]; + mip_swizzle_view_.label = [texture_ label]; texture_view_dirty_flags_ = TEXTURE_VIEW_NOT_DIRTY; } } @@ -180,29 +173,29 @@ id<MTLTexture> gpu::MTLTexture::get_metal_handle() this->ensure_baked(); /* Verify VBO texture shares same buffer. */ - if (this->resource_mode_ == MTL_TEXTURE_MODE_VBO) { + if (resource_mode_ == MTL_TEXTURE_MODE_VBO) { int r_offset = -1; /* TODO(Metal): Fetch buffer from MTLVertBuf when implemented. */ id<MTLBuffer> buf = nil; /*vert_buffer_->get_metal_buffer(&r_offset);*/ - BLI_assert(this->vert_buffer_mtl_ != nil); - BLI_assert(buf == this->vert_buffer_mtl_ && r_offset == this->vert_buffer_offset_); + BLI_assert(vert_buffer_mtl_ != nil); + BLI_assert(buf == vert_buffer_mtl_ && r_offset == vert_buffer_offset_); UNUSED_VARS(buf); UNUSED_VARS_NDEBUG(r_offset); } - if (this->is_baked_) { + if (is_baked_) { /* For explicit texture views, ensure we always return the texture view. */ - if (this->resource_mode_ == MTL_TEXTURE_MODE_TEXTURE_VIEW) { - BLI_assert(this->mip_swizzle_view_ && "Texture view should always have a valid handle."); + if (resource_mode_ == MTL_TEXTURE_MODE_TEXTURE_VIEW) { + BLI_assert(mip_swizzle_view_ && "Texture view should always have a valid handle."); } - if (this->mip_swizzle_view_ != nil || texture_view_dirty_flags_) { + if (mip_swizzle_view_ != nil || texture_view_dirty_flags_) { bake_mip_swizzle_view(); - return this->mip_swizzle_view_; + return mip_swizzle_view_; } - return this->texture_; + return texture_; } return nil; } @@ -214,36 +207,36 @@ id<MTLTexture> gpu::MTLTexture::get_metal_handle_base() this->ensure_baked(); /* For explicit texture views, always return the texture view. */ - if (this->resource_mode_ == MTL_TEXTURE_MODE_TEXTURE_VIEW) { - BLI_assert(this->mip_swizzle_view_ && "Texture view should always have a valid handle."); - if (this->mip_swizzle_view_ != nil || texture_view_dirty_flags_) { + if (resource_mode_ == MTL_TEXTURE_MODE_TEXTURE_VIEW) { + BLI_assert(mip_swizzle_view_ && "Texture view should always have a valid handle."); + if (mip_swizzle_view_ != nil || texture_view_dirty_flags_) { bake_mip_swizzle_view(); } - return this->mip_swizzle_view_; + return mip_swizzle_view_; } /* Return base handle. */ - if (this->is_baked_) { - return this->texture_; + if (is_baked_) { + return texture_; } return nil; } void gpu::MTLTexture::blit(id<MTLBlitCommandEncoder> blit_encoder, - unsigned int src_x_offset, - unsigned int src_y_offset, - unsigned int src_z_offset, - unsigned int src_slice, - unsigned int src_mip, + uint src_x_offset, + uint src_y_offset, + uint src_z_offset, + uint src_slice, + uint src_mip, gpu::MTLTexture *dest, - unsigned int dst_x_offset, - unsigned int dst_y_offset, - unsigned int dst_z_offset, - unsigned int dst_slice, - unsigned int dst_mip, - unsigned int width, - unsigned int height, - unsigned int depth) + uint dst_x_offset, + uint dst_y_offset, + uint dst_z_offset, + uint dst_slice, + uint dst_mip, + uint width, + uint height, + uint depth) { BLI_assert(this && dest); @@ -273,13 +266,13 @@ void gpu::MTLTexture::blit(id<MTLBlitCommandEncoder> blit_encoder, } void gpu::MTLTexture::blit(gpu::MTLTexture *dst, - unsigned int src_x_offset, - unsigned int src_y_offset, - unsigned int dst_x_offset, - unsigned int dst_y_offset, - unsigned int src_mip, - unsigned int dst_mip, - unsigned int dst_slice, + uint src_x_offset, + uint src_y_offset, + uint dst_x_offset, + uint dst_y_offset, + uint src_mip, + uint dst_mip, + uint dst_slice, int width, int height) { @@ -348,19 +341,19 @@ void gpu::MTLTexture::blit(gpu::MTLTexture *dst, } } -GPUFrameBuffer *gpu::MTLTexture::get_blit_framebuffer(unsigned int dst_slice, unsigned int dst_mip) +GPUFrameBuffer *gpu::MTLTexture::get_blit_framebuffer(uint dst_slice, uint dst_mip) { /* Check if layer has changed. */ bool update_attachments = false; - if (!this->blit_fb_) { - this->blit_fb_ = GPU_framebuffer_create("gpu_blit"); + if (!blit_fb_) { + blit_fb_ = GPU_framebuffer_create("gpu_blit"); update_attachments = true; } /* Check if current blit FB has the correct attachment properties. */ - if (this->blit_fb_) { - if (this->blit_fb_slice_ != dst_slice || this->blit_fb_mip_ != dst_mip) { + if (blit_fb_) { + if (blit_fb_slice_ != dst_slice || blit_fb_mip_ != dst_mip) { update_attachments = true; } } @@ -369,7 +362,7 @@ GPUFrameBuffer *gpu::MTLTexture::get_blit_framebuffer(unsigned int dst_slice, un if (format_flag_ & GPU_FORMAT_DEPTH || format_flag_ & GPU_FORMAT_STENCIL) { /* DEPTH TEX */ GPU_framebuffer_ensure_config( - &this->blit_fb_, + &blit_fb_, {GPU_ATTACHMENT_TEXTURE_LAYER_MIP(wrap(static_cast<Texture *>(this)), static_cast<int>(dst_slice), static_cast<int>(dst_mip)), @@ -378,18 +371,18 @@ GPUFrameBuffer *gpu::MTLTexture::get_blit_framebuffer(unsigned int dst_slice, un else { /* COLOR TEX */ GPU_framebuffer_ensure_config( - &this->blit_fb_, + &blit_fb_, {GPU_ATTACHMENT_NONE, GPU_ATTACHMENT_TEXTURE_LAYER_MIP(wrap(static_cast<Texture *>(this)), static_cast<int>(dst_slice), static_cast<int>(dst_mip))}); } - this->blit_fb_slice_ = dst_slice; - this->blit_fb_mip_ = dst_mip; + blit_fb_slice_ = dst_slice; + blit_fb_mip_ = dst_mip; } - BLI_assert(this->blit_fb_); - return this->blit_fb_; + BLI_assert(blit_fb_); + return blit_fb_; } MTLSamplerState gpu::MTLTexture::get_sampler_state() @@ -408,7 +401,7 @@ void gpu::MTLTexture::update_sub( BLI_assert(ctx); /* Do not update texture view. */ - BLI_assert(this->resource_mode_ != MTL_TEXTURE_MODE_TEXTURE_VIEW); + BLI_assert(resource_mode_ != MTL_TEXTURE_MODE_TEXTURE_VIEW); /* Ensure mipmaps. */ this->ensure_mipmaps(mip); @@ -418,16 +411,16 @@ void gpu::MTLTexture::update_sub( /* Safety checks. */ #if TRUST_NO_ONE - BLI_assert(mip >= this->mip_min_ && mip <= this->mip_max_); - BLI_assert(mip < this->texture_.mipmapLevelCount); - BLI_assert(this->texture_.mipmapLevelCount >= this->mip_max_); + BLI_assert(mip >= mip_min_ && mip <= mip_max_); + BLI_assert(mip < texture_.mipmapLevelCount); + BLI_assert(texture_.mipmapLevelCount >= mip_max_); #endif /* DEPTH FLAG - Depth formats cannot use direct BLIT - pass off to their own routine which will * do a depth-only render. */ - bool is_depth_format = (this->format_flag_ & GPU_FORMAT_DEPTH); + bool is_depth_format = (format_flag_ & GPU_FORMAT_DEPTH); if (is_depth_format) { - switch (this->type_) { + switch (type_) { case GPU_TEXTURE_2D: { update_sub_depth_2d(mip, offset, extent, type, data); @@ -444,7 +437,7 @@ void gpu::MTLTexture::update_sub( @autoreleasepool { /* Determine totalsize of INPUT Data. */ - int num_channels = to_component_len(this->format_); + int num_channels = to_component_len(format_); int input_bytes_per_pixel = num_channels * to_bytesize(type); int totalsize = 0; @@ -482,7 +475,7 @@ void gpu::MTLTexture::update_sub( BLI_assert(totalsize > 0); /* Determine expected destination data size. */ - MTLPixelFormat destination_format = gpu_texture_format_to_metal(this->format_); + MTLPixelFormat destination_format = gpu_texture_format_to_metal(format_); int expected_dst_bytes_per_pixel = get_mtl_format_bytesize(destination_format); int destination_num_channels = get_mtl_format_num_components(destination_format); int destination_totalsize = 0; @@ -517,21 +510,21 @@ void gpu::MTLTexture::update_sub( can_use_direct_blit = false; } -#if MTL_VALIDATION_CRASH_DEPTH_1_1_1_WA - if (this->type_ == GPU_TEXTURE_2D || this->type_ == GPU_TEXTURE_2D_ARRAY) { - /* Workaround for crash in validation layer when blitting to depth2D target with - * dimensions (1, 1, 1); */ - if (extent[0] == 1 && extent[1] == 1 && extent[2] == 1 && totalsize == 4) { - can_use_direct_blit = false; + if (is_depth_format) { + if (type_ == GPU_TEXTURE_2D || type_ == GPU_TEXTURE_2D_ARRAY) { + /* Workaround for crash in validation layer when blitting to depth2D target with + * dimensions (1, 1, 1); */ + if (extent[0] == 1 && extent[1] == 1 && extent[2] == 1 && totalsize == 4) { + can_use_direct_blit = false; + } } } -#endif - if (this->format_ == GPU_SRGB8_A8 && !can_use_direct_blit) { + if (format_ == GPU_SRGB8_A8 && !can_use_direct_blit) { MTL_LOG_WARNING( "SRGB data upload does not work correctly using compute upload. " "texname '%s'\n", - this->name_); + name_); } /* Safety Checks. */ @@ -573,35 +566,6 @@ void gpu::MTLTexture::update_sub( } } - /* Debug hook for performing GPU capture of routine. */ - bool DO_CAPTURE = false; -#if DEBUG_TEXTURE_UPDATE_CAPTURE == 1 - DO_CAPTURE = true; - if (DO_CAPTURE) { - MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager]; - MTLCaptureDescriptor *capture_descriptor = [[MTLCaptureDescriptor alloc] init]; - capture_descriptor.captureObject = ctx->device; - NSError *error; - if (![capture_manager startCaptureWithDescriptor:capture_descriptor error:&error]) { - NSString *error_str = [NSString stringWithFormat:@"%@", error]; - const char *error_c_str = [error_str UTF8String]; - MTL_LOG_ERROR("Failed to start capture. Error: %s\n", error_c_str); - } - } -#endif - - /* Fetch or Create command buffer. */ - id<MTLCommandBuffer> cmd_buffer = ctx->get_active_command_buffer(); - bool own_command_buffer = false; - if (cmd_buffer == nil || DO_CAPTURE) { - cmd_buffer = [ctx->queue commandBuffer]; - own_command_buffer = true; - } - else { - /* Finish graphics work. */ - ctx->end_render_pass(); - } - /* Prepare staging buffer for data. */ id<MTLBuffer> staging_buffer = nil; unsigned long long staging_buffer_offset = 0; @@ -611,11 +575,6 @@ void gpu::MTLTexture::update_sub( /* = ctx->get_memory_manager().scratch_buffer_allocate_range_aligned(totalsize, 256);*/ memcpy(allocation.host_ptr, data, totalsize); staging_buffer = allocation.metal_buffer; - if (own_command_buffer) { - if (allocation.requires_flush()) { - [staging_buffer didModifyRange:NSMakeRange(allocation.buffer_offset, allocation.size)]; - } - } staging_buffer_offset = allocation.buffer_offset; /* Common Properties. */ @@ -629,23 +588,23 @@ void gpu::MTLTexture::update_sub( return; } id<MTLTexture> texture_handle = ((compatible_write_format == destination_format)) ? - this->texture_ : - [this->texture_ + texture_ : + [texture_ newTextureViewWithPixelFormat:compatible_write_format]; - /* Prepare encoders */ + /* Prepare command encoders. */ id<MTLBlitCommandEncoder> blit_encoder = nil; id<MTLComputeCommandEncoder> compute_encoder = nil; if (can_use_direct_blit) { - blit_encoder = [cmd_buffer blitCommandEncoder]; + blit_encoder = ctx->main_command_buffer.ensure_begin_blit_encoder(); BLI_assert(blit_encoder != nil); } else { - compute_encoder = [cmd_buffer computeCommandEncoder]; + compute_encoder = ctx->main_command_buffer.ensure_begin_compute_encoder(); BLI_assert(compute_encoder != nil); } - switch (this->type_) { + switch (type_) { /* 1D */ case GPU_TEXTURE_1D: @@ -657,26 +616,26 @@ void gpu::MTLTexture::update_sub( extent[0] : ctx->pipeline_state.unpack_row_length); int bytes_per_image = bytes_per_row; - int max_array_index = ((this->type_ == GPU_TEXTURE_1D_ARRAY) ? extent[1] : 1); + int max_array_index = ((type_ == GPU_TEXTURE_1D_ARRAY) ? extent[1] : 1); for (int array_index = 0; array_index < max_array_index; array_index++) { int buffer_array_offset = staging_buffer_offset + (bytes_per_image * array_index); - [blit_encoder copyFromBuffer:staging_buffer - sourceOffset:buffer_array_offset - sourceBytesPerRow:bytes_per_row - sourceBytesPerImage:bytes_per_image - sourceSize:MTLSizeMake(extent[0], 1, 1) - toTexture:texture_handle - destinationSlice:((this->type_ == GPU_TEXTURE_1D_ARRAY) ? - (array_index + offset[1]) : - 0) - destinationLevel:mip - destinationOrigin:MTLOriginMake(offset[0], 0, 0)]; + [blit_encoder + copyFromBuffer:staging_buffer + sourceOffset:buffer_array_offset + sourceBytesPerRow:bytes_per_row + sourceBytesPerImage:bytes_per_image + sourceSize:MTLSizeMake(extent[0], 1, 1) + toTexture:texture_handle + destinationSlice:((type_ == GPU_TEXTURE_1D_ARRAY) ? (array_index + offset[1]) : + 0) + destinationLevel:mip + destinationOrigin:MTLOriginMake(offset[0], 0, 0)]; } } else { /* Use Compute Based update. */ - if (this->type_ == GPU_TEXTURE_1D) { + if (type_ == GPU_TEXTURE_1D) { id<MTLComputePipelineState> pso = texture_update_1d_get_kernel( compute_specialisation_kernel); TextureUpdateParams params = {mip, @@ -693,7 +652,7 @@ void gpu::MTLTexture::update_sub( dispatchThreads:MTLSizeMake(extent[0], 1, 1) /* Width, Height, Layer */ threadsPerThreadgroup:MTLSizeMake(64, 1, 1)]; } - else if (this->type_ == GPU_TEXTURE_1D_ARRAY) { + else if (type_ == GPU_TEXTURE_1D_ARRAY) { id<MTLComputePipelineState> pso = texture_update_1d_array_get_kernel( compute_specialisation_kernel); TextureUpdateParams params = {mip, @@ -725,14 +684,14 @@ void gpu::MTLTexture::update_sub( int bytes_per_image = bytes_per_row * extent[1]; int texture_array_relative_offset = 0; - int base_slice = (this->type_ == GPU_TEXTURE_2D_ARRAY) ? offset[2] : 0; - int final_slice = base_slice + ((this->type_ == GPU_TEXTURE_2D_ARRAY) ? extent[2] : 1); + int base_slice = (type_ == GPU_TEXTURE_2D_ARRAY) ? offset[2] : 0; + int final_slice = base_slice + ((type_ == GPU_TEXTURE_2D_ARRAY) ? extent[2] : 1); for (int array_slice = base_slice; array_slice < final_slice; array_slice++) { if (array_slice > 0) { - BLI_assert(this->type_ == GPU_TEXTURE_2D_ARRAY); - BLI_assert(array_slice < this->d_); + BLI_assert(type_ == GPU_TEXTURE_2D_ARRAY); + BLI_assert(array_slice < d_); } [blit_encoder copyFromBuffer:staging_buffer @@ -750,7 +709,7 @@ void gpu::MTLTexture::update_sub( } else { /* Use Compute texture update. */ - if (this->type_ == GPU_TEXTURE_2D) { + if (type_ == GPU_TEXTURE_2D) { id<MTLComputePipelineState> pso = texture_update_2d_get_kernel( compute_specialisation_kernel); TextureUpdateParams params = {mip, @@ -768,7 +727,7 @@ void gpu::MTLTexture::update_sub( extent[0], extent[1], 1) /* Width, Height, Layer */ threadsPerThreadgroup:MTLSizeMake(8, 8, 1)]; } - else if (this->type_ == GPU_TEXTURE_2D_ARRAY) { + else if (type_ == GPU_TEXTURE_2D_ARRAY) { id<MTLComputePipelineState> pso = texture_update_2d_array_get_kernel( compute_specialisation_kernel); TextureUpdateParams params = {mip, @@ -918,35 +877,15 @@ void gpu::MTLTexture::update_sub( if (texture_.storageMode == MTLStorageModeManaged) { [blit_encoder synchronizeResource:texture_buffer_]; } - - /* End Encoding. */ - [blit_encoder endEncoding]; } else { - - /* End Encoding. */ - [compute_encoder endEncoding]; - /* Textures which use MTLStorageModeManaged need to have updated contents * synced back to CPU to avoid an automatic flush overwriting contents. */ if (texture_.storageMode == MTLStorageModeManaged) { - blit_encoder = [cmd_buffer blitCommandEncoder]; + blit_encoder = ctx->main_command_buffer.ensure_begin_blit_encoder(); [blit_encoder synchronizeResource:texture_buffer_]; - [blit_encoder endEncoding]; } } - - if (own_command_buffer) { - [cmd_buffer commit]; - } - -#if DEBUG_TEXTURE_UPDATE_CAPTURE == 1 - if (DO_CAPTURE) { - [cmd_buffer waitUntilCompleted]; - MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager]; - [capture_manager stopCapture]; - } -#endif } } @@ -954,12 +893,12 @@ void gpu::MTLTexture::ensure_mipmaps(int miplvl) { /* Do not update texture view. */ - BLI_assert(this->resource_mode_ != MTL_TEXTURE_MODE_TEXTURE_VIEW); + BLI_assert(resource_mode_ != MTL_TEXTURE_MODE_TEXTURE_VIEW); /* Clamp level to maximum. */ - int effective_h = (this->type_ == GPU_TEXTURE_1D_ARRAY) ? 0 : this->h_; - int effective_d = (this->type_ != GPU_TEXTURE_3D) ? 0 : this->d_; - int max_dimension = max_iii(this->w_, effective_h, effective_d); + int effective_h = (type_ == GPU_TEXTURE_1D_ARRAY) ? 0 : h_; + int effective_d = (type_ != GPU_TEXTURE_3D) ? 0 : d_; + int max_dimension = max_iii(w_, effective_h, effective_d); int max_miplvl = floor(log2(max_dimension)); miplvl = min_ii(max_miplvl, miplvl); @@ -968,8 +907,8 @@ void gpu::MTLTexture::ensure_mipmaps(int miplvl) mipmaps_ = miplvl; /* Check if baked. */ - if (this->is_baked_ && mipmaps_ > mtl_max_mips_) { - this->is_dirty_ = true; + if (is_baked_ && mipmaps_ > mtl_max_mips_) { + is_dirty_ = true; MTL_LOG_WARNING("Texture requires regenerating due to increase in mip-count\n"); } } @@ -993,44 +932,29 @@ void gpu::MTLTexture::generate_mipmap(void) /* Ensure texture is baked. */ this->ensure_baked(); - BLI_assert(this->is_baked_ && this->texture_ && "MTLTexture is not valid"); + BLI_assert(is_baked_ && texture_ && "MTLTexture is not valid"); - if (this->mipmaps_ == 1 || this->mtl_max_mips_ == 1) { + if (mipmaps_ == 1 || mtl_max_mips_ == 1) { MTL_LOG_WARNING("Call to generate mipmaps on texture with 'mipmaps_=1\n'"); return; } /* Verify if we can perform mipmap generation. */ - if (this->format_ == GPU_DEPTH_COMPONENT32F || this->format_ == GPU_DEPTH_COMPONENT24 || - this->format_ == GPU_DEPTH_COMPONENT16 || this->format_ == GPU_DEPTH32F_STENCIL8 || - this->format_ == GPU_DEPTH24_STENCIL8) { + if (format_ == GPU_DEPTH_COMPONENT32F || format_ == GPU_DEPTH_COMPONENT24 || + format_ == GPU_DEPTH_COMPONENT16 || format_ == GPU_DEPTH32F_STENCIL8 || + format_ == GPU_DEPTH24_STENCIL8) { MTL_LOG_WARNING("Cannot generate mipmaps for textures using DEPTH formats\n"); return; } @autoreleasepool { - id<MTLCommandBuffer> cmd_buffer = ctx->get_active_command_buffer(); - bool own_command_buffer = false; - if (cmd_buffer == nil) { - cmd_buffer = [ctx->queue commandBuffer]; - own_command_buffer = true; - } - else { - /* End active graphics work. */ - ctx->end_render_pass(); - } - - id<MTLBlitCommandEncoder> enc = [cmd_buffer blitCommandEncoder]; -#if MTL_DEBUG_COMMAND_BUFFER_EXECUTION - [enc insertDebugSignpost:@"Generate MipMaps"]; -#endif - [enc generateMipmapsForTexture:this->texture_]; - [enc endEncoding]; - - if (own_command_buffer) { - [cmd_buffer commit]; + /* Fetch active BlitCommandEncoder. */ + id<MTLBlitCommandEncoder> enc = ctx->main_command_buffer.ensure_begin_blit_encoder(); + if (G.debug & G_DEBUG_GPU) { + [enc insertDebugSignpost:@"Generate MipMaps"]; } + [enc generateMipmapsForTexture:texture_]; } return; } @@ -1055,13 +979,8 @@ void gpu::MTLTexture::copy_to(Texture *dst) this->ensure_baked(); @autoreleasepool { - /* End render pass. */ - ctx->end_render_pass(); - /* Setup blit encoder. */ - id<MTLCommandBuffer> cmd_buffer = ctx->get_active_command_buffer(); - BLI_assert(cmd_buffer != nil); - id<MTLBlitCommandEncoder> blit_encoder = [cmd_buffer blitCommandEncoder]; + id<MTLBlitCommandEncoder> blit_encoder = ctx->main_command_buffer.ensure_begin_blit_encoder(); BLI_assert(blit_encoder != nil); /* TODO(Metal): Consider supporting multiple mip levels IF the GL implementation @@ -1077,7 +996,7 @@ void gpu::MTLTexture::copy_to(Texture *dst) case GPU_TEXTURE_CUBE_ARRAY: case GPU_TEXTURE_3D: { /* Do full texture copy for 3D textures */ - BLI_assert(mt_dst->d_ == this->d_); + BLI_assert(mt_dst->d_ == d_); [blit_encoder copyFromTexture:this->get_metal_handle_base() toTexture:mt_dst->get_metal_handle_base()]; } break; @@ -1100,9 +1019,6 @@ void gpu::MTLTexture::copy_to(Texture *dst) extent[2]); } break; } - - /* End encoding */ - [blit_encoder endEncoding]; } } @@ -1144,8 +1060,8 @@ static MTLTextureSwizzle swizzle_to_mtl(const char swizzle) void gpu::MTLTexture::swizzle_set(const char swizzle_mask[4]) { - if (memcmp(this->tex_swizzle_mask_, swizzle_mask, 4) != 0) { - memcpy(this->tex_swizzle_mask_, swizzle_mask, 4); + if (memcmp(tex_swizzle_mask_, swizzle_mask, 4) != 0) { + memcpy(tex_swizzle_mask_, swizzle_mask, 4); /* Creating the swizzle mask and flagging as dirty if changed. */ MTLTextureSwizzleChannels new_swizzle_mask = MTLTextureSwizzleChannelsMake( @@ -1154,8 +1070,8 @@ void gpu::MTLTexture::swizzle_set(const char swizzle_mask[4]) swizzle_to_mtl(swizzle_mask[2]), swizzle_to_mtl(swizzle_mask[3])); - this->mtl_swizzle_mask_ = new_swizzle_mask; - this->texture_view_dirty_flags_ |= TEXTURE_VIEW_SWIZZLE_DIRTY; + mtl_swizzle_mask_ = new_swizzle_mask; + texture_view_dirty_flags_ |= TEXTURE_VIEW_SWIZZLE_DIRTY; } } @@ -1172,25 +1088,24 @@ void gpu::MTLTexture::mip_range_set(int min, int max) * * TODO(Metal): Add texture initialization flag to determine whether mipmaps are used * or not. Will be important for saving memory for big textures. */ - this->mip_min_ = min; - this->mip_max_ = max; + mip_min_ = min; + mip_max_ = max; - if ((this->type_ == GPU_TEXTURE_1D || this->type_ == GPU_TEXTURE_1D_ARRAY || - this->type_ == GPU_TEXTURE_BUFFER) && + if ((type_ == GPU_TEXTURE_1D || type_ == GPU_TEXTURE_1D_ARRAY || type_ == GPU_TEXTURE_BUFFER) && max > 1) { MTL_LOG_ERROR( " MTLTexture of type TEXTURE_1D_ARRAY or TEXTURE_BUFFER cannot have a mipcount " "greater than 1\n"); - this->mip_min_ = 0; - this->mip_max_ = 0; - this->mipmaps_ = 0; + mip_min_ = 0; + mip_max_ = 0; + mipmaps_ = 0; BLI_assert(false); } /* Mip range for texture view. */ - this->mip_texture_base_level_ = this->mip_min_; - this->mip_texture_max_level_ = this->mip_max_; + mip_texture_base_level_ = mip_min_; + mip_texture_max_level_ = mip_max_; texture_view_dirty_flags_ |= TEXTURE_VIEW_MIP_DIRTY; } @@ -1199,7 +1114,7 @@ void *gpu::MTLTexture::read(int mip, eGPUDataFormat type) /* Prepare Array for return data. */ BLI_assert(!(format_flag_ & GPU_FORMAT_COMPRESSED)); BLI_assert(mip <= mipmaps_); - BLI_assert(validate_data_format_mtl(this->format_, type)); + BLI_assert(validate_data_format_mtl(format_, type)); /* NOTE: mip_size_get() won't override any dimension that is equal to 0. */ int extent[3] = {1, 1, 1}; @@ -1208,12 +1123,12 @@ void *gpu::MTLTexture::read(int mip, eGPUDataFormat type) size_t sample_len = extent[0] * extent[1] * extent[2]; size_t sample_size = to_bytesize(format_, type); size_t texture_size = sample_len * sample_size; - int num_channels = to_component_len(this->format_); + int num_channels = to_component_len(format_); void *data = MEM_mallocN(texture_size + 8, "GPU_texture_read"); /* Ensure texture is baked. */ - if (this->is_baked_) { + if (is_baked_) { this->read_internal( mip, 0, 0, 0, extent[0], extent[1], extent[2], type, num_channels, texture_size + 8, data); } @@ -1239,7 +1154,7 @@ void gpu::MTLTexture::read_internal(int mip, void *r_data) { /* Verify textures are baked. */ - if (!this->is_baked_) { + if (!is_baked_) { MTL_LOG_WARNING("gpu::MTLTexture::read_internal - Trying to read from a non-baked texture!\n"); return; } @@ -1248,14 +1163,14 @@ void gpu::MTLTexture::read_internal(int mip, BLI_assert(ctx); /* Calculate Desired output size. */ - int num_channels = to_component_len(this->format_); + int num_channels = to_component_len(format_); BLI_assert(num_output_components <= num_channels); - unsigned int desired_output_bpp = num_output_components * to_bytesize(desired_output_format); + uint desired_output_bpp = num_output_components * to_bytesize(desired_output_format); /* Calculate Metal data output for trivial copy. */ - unsigned int image_bpp = get_mtl_format_bytesize(this->texture_.pixelFormat); - unsigned int image_components = get_mtl_format_num_components(this->texture_.pixelFormat); - bool is_depth_format = (this->format_flag_ & GPU_FORMAT_DEPTH); + uint image_bpp = get_mtl_format_bytesize(texture_.pixelFormat); + uint image_components = get_mtl_format_num_components(texture_.pixelFormat); + bool is_depth_format = (format_flag_ & GPU_FORMAT_DEPTH); /* Verify if we need to use compute read. */ eGPUDataFormat data_format = to_mtl_internal_data_format(this->format_get()); @@ -1272,12 +1187,12 @@ void gpu::MTLTexture::read_internal(int mip, BLI_assert(num_output_components == 1); BLI_assert(image_components == 1); BLI_assert(data_format == GPU_DATA_FLOAT || data_format == GPU_DATA_UINT_24_8); - BLI_assert(validate_data_format_mtl(this->format_, data_format)); + BLI_assert(validate_data_format_mtl(format_, data_format)); } /* SPECIAL Workaround for R11G11B10 textures requesting a read using: GPU_DATA_10_11_11_REV. */ if (desired_output_format == GPU_DATA_10_11_11_REV) { - BLI_assert(this->format_ == GPU_R11F_G11F_B10F); + BLI_assert(format_ == GPU_R11F_G11F_B10F); /* override parameters - we'll be able to use simple copy, as bpp will match at 4 bytes. */ image_bpp = sizeof(int); @@ -1291,9 +1206,9 @@ void gpu::MTLTexture::read_internal(int mip, } /* Determine size of output data. */ - unsigned int bytes_per_row = desired_output_bpp * width; - unsigned int bytes_per_image = bytes_per_row * height; - unsigned int total_bytes = bytes_per_image * depth; + uint bytes_per_row = desired_output_bpp * width; + uint bytes_per_image = bytes_per_row * height; + uint total_bytes = bytes_per_image * depth; if (can_use_simple_read) { /* DEBUG check that if direct copy is being used, then both the expected output size matches @@ -1307,7 +1222,7 @@ void gpu::MTLTexture::read_internal(int mip, /* Fetch allocation from scratch buffer. */ id<MTLBuffer> destination_buffer = nil; - unsigned int destination_offset = 0; + uint destination_offset = 0; void *destination_buffer_host_ptr = nullptr; /* TODO(Metal): Optimize buffer allocation. */ @@ -1348,53 +1263,25 @@ void gpu::MTLTexture::read_internal(int mip, bool copy_successful = false; @autoreleasepool { - bool DO_CAPTURE = false; -#if DEBUG_TEXTURE_READ_CAPTURE == 1 - DO_CAPTURE = true; - if (DO_CAPTURE) { - MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager]; - MTLCaptureDescriptor *capture_descriptor = [[MTLCaptureDescriptor alloc] init]; - capture_descriptor.captureObject = ctx->device; - NSError *error; - if (![capture_manager startCaptureWithDescriptor:capture_descriptor error:&error]) { - NSString *error_str = [NSString stringWithFormat:@"%@", error]; - const char *error_c_str = [error_str UTF8String]; - MTL_LOG_ERROR("Failed to start capture. Error: %s\n", error_c_str); - } - } -#endif - /* TODO(Metal): Verify whether we need some form of barrier here to ensure reads * happen after work with associated texture is finished. */ GPU_finish(); - /* Fetch or Create command buffer. */ - id<MTLCommandBuffer> cmd_buffer = ctx->get_active_command_buffer(); - bool own_command_buffer = false; - if (cmd_buffer == nil || DO_CAPTURE || true) { - cmd_buffer = [ctx->queue commandBuffer]; - own_command_buffer = true; - } - else { - /* End any graphics workloads. */ - ctx->end_render_pass(); - } - /* Texture View for SRGB special case. */ - id<MTLTexture> read_texture = this->texture_; - if (this->format_ == GPU_SRGB8_A8) { - read_texture = [this->texture_ newTextureViewWithPixelFormat:MTLPixelFormatRGBA8Unorm]; + id<MTLTexture> read_texture = texture_; + if (format_ == GPU_SRGB8_A8) { + read_texture = [texture_ newTextureViewWithPixelFormat:MTLPixelFormatRGBA8Unorm]; } /* Perform per-texture type read. */ - switch (this->type_) { + switch (type_) { case GPU_TEXTURE_2D: { if (can_use_simple_read) { /* Use Blit Encoder READ. */ - id<MTLBlitCommandEncoder> enc = [cmd_buffer blitCommandEncoder]; -#if MTL_DEBUG_COMMAND_BUFFER_EXECUTION - [enc insertDebugSignpost:@"GPUTextureRead"]; -#endif + id<MTLBlitCommandEncoder> enc = ctx->main_command_buffer.ensure_begin_blit_encoder(); + if (G.debug & G_DEBUG_GPU) { + [enc insertDebugSignpost:@"GPUTextureRead"]; + } [enc copyFromTexture:read_texture sourceSlice:0 sourceLevel:mip @@ -1405,13 +1292,13 @@ void gpu::MTLTexture::read_internal(int mip, destinationBytesPerRow:bytes_per_row destinationBytesPerImage:bytes_per_image]; [enc synchronizeResource:destination_buffer]; - [enc endEncoding]; copy_successful = true; } else { /* Use Compute READ. */ - id<MTLComputeCommandEncoder> compute_encoder = [cmd_buffer computeCommandEncoder]; + id<MTLComputeCommandEncoder> compute_encoder = + ctx->main_command_buffer.ensure_begin_compute_encoder(); id<MTLComputePipelineState> pso = texture_read_2d_get_kernel( compute_specialisation_kernel); TextureReadParams params = { @@ -1425,15 +1312,13 @@ void gpu::MTLTexture::read_internal(int mip, [compute_encoder setTexture:read_texture atIndex:0]; [compute_encoder dispatchThreads:MTLSizeMake(width, height, 1) /* Width, Height, Layer */ threadsPerThreadgroup:MTLSizeMake(8, 8, 1)]; - [compute_encoder endEncoding]; /* Use Blit encoder to synchronize results back to CPU. */ - id<MTLBlitCommandEncoder> enc = [cmd_buffer blitCommandEncoder]; -#if MTL_DEBUG_COMMAND_BUFFER_EXECUTION - [enc insertDebugSignpost:@"GPUTextureRead-syncResource"]; -#endif + id<MTLBlitCommandEncoder> enc = ctx->main_command_buffer.ensure_begin_blit_encoder(); + if (G.debug & G_DEBUG_GPU) { + [enc insertDebugSignpost:@"GPUTextureRead-syncResource"]; + } [enc synchronizeResource:destination_buffer]; - [enc endEncoding]; copy_successful = true; } } break; @@ -1441,10 +1326,10 @@ void gpu::MTLTexture::read_internal(int mip, case GPU_TEXTURE_2D_ARRAY: { if (can_use_simple_read) { /* Use Blit Encoder READ. */ - id<MTLBlitCommandEncoder> enc = [cmd_buffer blitCommandEncoder]; -#if MTL_DEBUG_COMMAND_BUFFER_EXECUTION - [enc insertDebugSignpost:@"GPUTextureRead"]; -#endif + id<MTLBlitCommandEncoder> enc = ctx->main_command_buffer.ensure_begin_blit_encoder(); + if (G.debug & G_DEBUG_GPU) { + [enc insertDebugSignpost:@"GPUTextureRead"]; + } int base_slice = z_off; int final_slice = base_slice + depth; int texture_array_relative_offset = 0; @@ -1463,13 +1348,13 @@ void gpu::MTLTexture::read_internal(int mip, texture_array_relative_offset += bytes_per_image; } - [enc endEncoding]; copy_successful = true; } else { /* Use Compute READ */ - id<MTLComputeCommandEncoder> compute_encoder = [cmd_buffer computeCommandEncoder]; + id<MTLComputeCommandEncoder> compute_encoder = + ctx->main_command_buffer.ensure_begin_compute_encoder(); id<MTLComputePipelineState> pso = texture_read_2d_array_get_kernel( compute_specialisation_kernel); TextureReadParams params = { @@ -1484,25 +1369,23 @@ void gpu::MTLTexture::read_internal(int mip, [compute_encoder dispatchThreads:MTLSizeMake(width, height, depth) /* Width, Height, Layer */ threadsPerThreadgroup:MTLSizeMake(8, 8, 1)]; - [compute_encoder endEncoding]; /* Use Blit encoder to synchronize results back to CPU. */ - id<MTLBlitCommandEncoder> enc = [cmd_buffer blitCommandEncoder]; -#if MTL_DEBUG_COMMAND_BUFFER_EXECUTION - [enc insertDebugSignpost:@"GPUTextureRead-syncResource"]; -#endif + id<MTLBlitCommandEncoder> enc = ctx->main_command_buffer.ensure_begin_blit_encoder(); + if (G.debug & G_DEBUG_GPU) { + [enc insertDebugSignpost:@"GPUTextureRead-syncResource"]; + } [enc synchronizeResource:destination_buffer]; - [enc endEncoding]; copy_successful = true; } } break; case GPU_TEXTURE_CUBE_ARRAY: { if (can_use_simple_read) { - id<MTLBlitCommandEncoder> enc = [cmd_buffer blitCommandEncoder]; -#if MTL_DEBUG_COMMAND_BUFFER_EXECUTION - [enc insertDebugSignpost:@"GPUTextureRead"]; -#endif + id<MTLBlitCommandEncoder> enc = ctx->main_command_buffer.ensure_begin_blit_encoder(); + if (G.debug & G_DEBUG_GPU) { + [enc insertDebugSignpost:@"GPUTextureRead"]; + } int base_slice = z_off; int final_slice = base_slice + depth; int texture_array_relative_offset = 0; @@ -1522,7 +1405,6 @@ void gpu::MTLTexture::read_internal(int mip, texture_array_relative_offset += bytes_per_image; } MTL_LOG_INFO("Copying texture data to buffer GPU_TEXTURE_CUBE_ARRAY\n"); - [enc endEncoding]; copy_successful = true; } else { @@ -1534,27 +1416,13 @@ void gpu::MTLTexture::read_internal(int mip, MTL_LOG_WARNING( "[Warning] gpu::MTLTexture::read_internal simple-copy not yet supported for texture " "type: %d\n", - (int)this->type_); + (int)type_); break; } if (copy_successful) { - /* Ensure GPU copy from texture to host-accessible buffer is complete. */ - if (own_command_buffer) { - [cmd_buffer commit]; - [cmd_buffer waitUntilCompleted]; - } - else { - /* Ensure GPU copy commands have completed. */ - GPU_finish(); - } - -#if DEBUG_TEXTURE_READ_CAPTURE == 1 - if (DO_CAPTURE) { - MTLCaptureManager *capture_manager = [MTLCaptureManager sharedCaptureManager]; - [capture_manager stopCapture]; - } -#endif + /* Ensure GPU copy commands have completed. */ + GPU_finish(); /* Copy data from Shared Memory into ptr. */ memcpy(r_data, destination_buffer_host_ptr, total_bytes); @@ -1583,9 +1451,9 @@ uint gpu::MTLTexture::gl_bindcode_get(void) const bool gpu::MTLTexture::init_internal(void) { - if (this->format_ == GPU_DEPTH24_STENCIL8) { + if (format_ == GPU_DEPTH24_STENCIL8) { /* Apple Silicon requires GPU_DEPTH32F_STENCIL8 instead of GPU_DEPTH24_STENCIL8. */ - this->format_ = GPU_DEPTH32F_STENCIL8; + format_ = GPU_DEPTH32F_STENCIL8; } this->prepare_internal(); @@ -1609,20 +1477,20 @@ bool gpu::MTLTexture::init_internal(const GPUTexture *src, int mip_offset, int l this->prepare_internal(); /* Flag as using texture view. */ - this->resource_mode_ = MTL_TEXTURE_MODE_TEXTURE_VIEW; - this->source_texture_ = src; - this->mip_texture_base_level_ = mip_offset; - this->mip_texture_base_layer_ = layer_offset; + resource_mode_ = MTL_TEXTURE_MODE_TEXTURE_VIEW; + source_texture_ = src; + mip_texture_base_level_ = mip_offset; + mip_texture_base_layer_ = layer_offset; /* Assign texture as view. */ const gpu::MTLTexture *mtltex = static_cast<const gpu::MTLTexture *>(unwrap(src)); - this->texture_ = mtltex->texture_; - BLI_assert(this->texture_); - [this->texture_ retain]; + texture_ = mtltex->texture_; + BLI_assert(texture_); + [texture_ retain]; /* Flag texture as baked -- we do not need explicit initialization. */ - this->is_baked_ = true; - this->is_dirty_ = false; + is_baked_ = true; + is_dirty_ = false; /* Bake mip swizzle view. */ bake_mip_swizzle_view(); @@ -1637,7 +1505,7 @@ bool gpu::MTLTexture::init_internal(const GPUTexture *src, int mip_offset, int l bool gpu::MTLTexture::texture_is_baked() { - return this->is_baked_; + return is_baked_; } /* Prepare texture parameters after initialization, but before baking. */ @@ -1645,22 +1513,21 @@ void gpu::MTLTexture::prepare_internal() { /* Derive implicit usage flags for Depth/Stencil attachments. */ - if (this->format_flag_ & GPU_FORMAT_DEPTH || this->format_flag_ & GPU_FORMAT_STENCIL) { - this->gpu_image_usage_flags_ |= GPU_TEXTURE_USAGE_ATTACHMENT; + if (format_flag_ & GPU_FORMAT_DEPTH || format_flag_ & GPU_FORMAT_STENCIL) { + gpu_image_usage_flags_ |= GPU_TEXTURE_USAGE_ATTACHMENT; } /* Derive maximum number of mip levels by default. * TODO(Metal): This can be removed if max mip counts are specified upfront. */ - if (this->type_ == GPU_TEXTURE_1D || this->type_ == GPU_TEXTURE_1D_ARRAY || - this->type_ == GPU_TEXTURE_BUFFER) { - this->mtl_max_mips_ = 1; + if (type_ == GPU_TEXTURE_1D || type_ == GPU_TEXTURE_1D_ARRAY || type_ == GPU_TEXTURE_BUFFER) { + mtl_max_mips_ = 1; } else { - int effective_h = (this->type_ == GPU_TEXTURE_1D_ARRAY) ? 0 : this->h_; - int effective_d = (this->type_ != GPU_TEXTURE_3D) ? 0 : this->d_; - int max_dimension = max_iii(this->w_, effective_h, effective_d); + int effective_h = (type_ == GPU_TEXTURE_1D_ARRAY) ? 0 : h_; + int effective_d = (type_ != GPU_TEXTURE_3D) ? 0 : d_; + int max_dimension = max_iii(w_, effective_h, effective_d); int max_miplvl = max_ii(floor(log2(max_dimension)) + 1, 1); - this->mtl_max_mips_ = max_miplvl; + mtl_max_mips_ = max_miplvl; } } @@ -1669,101 +1536,91 @@ void gpu::MTLTexture::ensure_baked() /* If properties have changed, re-bake. */ bool copy_previous_contents = false; - if (this->is_baked_ && this->is_dirty_) { + if (is_baked_ && is_dirty_) { copy_previous_contents = true; - id<MTLTexture> previous_texture = this->texture_; + id<MTLTexture> previous_texture = texture_; [previous_texture retain]; this->reset(); } - if (!this->is_baked_) { + if (!is_baked_) { MTLContext *ctx = static_cast<MTLContext *>(unwrap(GPU_context_active_get())); BLI_assert(ctx); /* Ensure texture mode is valid. */ - BLI_assert(this->resource_mode_ != MTL_TEXTURE_MODE_EXTERNAL); - BLI_assert(this->resource_mode_ != MTL_TEXTURE_MODE_TEXTURE_VIEW); - BLI_assert(this->resource_mode_ != MTL_TEXTURE_MODE_VBO); + BLI_assert(resource_mode_ != MTL_TEXTURE_MODE_EXTERNAL); + BLI_assert(resource_mode_ != MTL_TEXTURE_MODE_TEXTURE_VIEW); + BLI_assert(resource_mode_ != MTL_TEXTURE_MODE_VBO); /* Format and mip levels (TODO(Metal): Optimize mipmaps counts, specify up-front). */ - MTLPixelFormat mtl_format = gpu_texture_format_to_metal(this->format_); + MTLPixelFormat mtl_format = gpu_texture_format_to_metal(format_); /* Create texture descriptor. */ - switch (this->type_) { + switch (type_) { /* 1D */ case GPU_TEXTURE_1D: case GPU_TEXTURE_1D_ARRAY: { - BLI_assert(this->w_ > 0); - this->texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; - this->texture_descriptor_.pixelFormat = mtl_format; - this->texture_descriptor_.textureType = (this->type_ == GPU_TEXTURE_1D_ARRAY) ? - MTLTextureType1DArray : - MTLTextureType1D; - this->texture_descriptor_.width = this->w_; - this->texture_descriptor_.height = 1; - this->texture_descriptor_.depth = 1; - this->texture_descriptor_.arrayLength = (this->type_ == GPU_TEXTURE_1D_ARRAY) ? this->h_ : - 1; - this->texture_descriptor_.mipmapLevelCount = (this->mtl_max_mips_ > 0) ? - this->mtl_max_mips_ : - 1; - this->texture_descriptor_.usage = + BLI_assert(w_ > 0); + texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; + texture_descriptor_.pixelFormat = mtl_format; + texture_descriptor_.textureType = (type_ == GPU_TEXTURE_1D_ARRAY) ? MTLTextureType1DArray : + MTLTextureType1D; + texture_descriptor_.width = w_; + texture_descriptor_.height = 1; + texture_descriptor_.depth = 1; + texture_descriptor_.arrayLength = (type_ == GPU_TEXTURE_1D_ARRAY) ? h_ : 1; + texture_descriptor_.mipmapLevelCount = (mtl_max_mips_ > 0) ? mtl_max_mips_ : 1; + texture_descriptor_.usage = MTLTextureUsageRenderTarget | MTLTextureUsageShaderRead | MTLTextureUsageShaderWrite | MTLTextureUsagePixelFormatView; /* TODO(Metal): Optimize usage flags. */ - this->texture_descriptor_.storageMode = MTLStorageModePrivate; - this->texture_descriptor_.sampleCount = 1; - this->texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; - this->texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; + texture_descriptor_.storageMode = MTLStorageModePrivate; + texture_descriptor_.sampleCount = 1; + texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; + texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; } break; /* 2D */ case GPU_TEXTURE_2D: case GPU_TEXTURE_2D_ARRAY: { - BLI_assert(this->w_ > 0 && this->h_ > 0); - this->texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; - this->texture_descriptor_.pixelFormat = mtl_format; - this->texture_descriptor_.textureType = (this->type_ == GPU_TEXTURE_2D_ARRAY) ? - MTLTextureType2DArray : - MTLTextureType2D; - this->texture_descriptor_.width = this->w_; - this->texture_descriptor_.height = this->h_; - this->texture_descriptor_.depth = 1; - this->texture_descriptor_.arrayLength = (this->type_ == GPU_TEXTURE_2D_ARRAY) ? this->d_ : - 1; - this->texture_descriptor_.mipmapLevelCount = (this->mtl_max_mips_ > 0) ? - this->mtl_max_mips_ : - 1; - this->texture_descriptor_.usage = + BLI_assert(w_ > 0 && h_ > 0); + texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; + texture_descriptor_.pixelFormat = mtl_format; + texture_descriptor_.textureType = (type_ == GPU_TEXTURE_2D_ARRAY) ? MTLTextureType2DArray : + MTLTextureType2D; + texture_descriptor_.width = w_; + texture_descriptor_.height = h_; + texture_descriptor_.depth = 1; + texture_descriptor_.arrayLength = (type_ == GPU_TEXTURE_2D_ARRAY) ? d_ : 1; + texture_descriptor_.mipmapLevelCount = (mtl_max_mips_ > 0) ? mtl_max_mips_ : 1; + texture_descriptor_.usage = MTLTextureUsageRenderTarget | MTLTextureUsageShaderRead | MTLTextureUsageShaderWrite | MTLTextureUsagePixelFormatView; /* TODO(Metal): Optimize usage flags. */ - this->texture_descriptor_.storageMode = MTLStorageModePrivate; - this->texture_descriptor_.sampleCount = 1; - this->texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; - this->texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; + texture_descriptor_.storageMode = MTLStorageModePrivate; + texture_descriptor_.sampleCount = 1; + texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; + texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; } break; /* 3D */ case GPU_TEXTURE_3D: { - BLI_assert(this->w_ > 0 && this->h_ > 0 && this->d_ > 0); - this->texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; - this->texture_descriptor_.pixelFormat = mtl_format; - this->texture_descriptor_.textureType = MTLTextureType3D; - this->texture_descriptor_.width = this->w_; - this->texture_descriptor_.height = this->h_; - this->texture_descriptor_.depth = this->d_; - this->texture_descriptor_.arrayLength = 1; - this->texture_descriptor_.mipmapLevelCount = (this->mtl_max_mips_ > 0) ? - this->mtl_max_mips_ : - 1; - this->texture_descriptor_.usage = + BLI_assert(w_ > 0 && h_ > 0 && d_ > 0); + texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; + texture_descriptor_.pixelFormat = mtl_format; + texture_descriptor_.textureType = MTLTextureType3D; + texture_descriptor_.width = w_; + texture_descriptor_.height = h_; + texture_descriptor_.depth = d_; + texture_descriptor_.arrayLength = 1; + texture_descriptor_.mipmapLevelCount = (mtl_max_mips_ > 0) ? mtl_max_mips_ : 1; + texture_descriptor_.usage = MTLTextureUsageRenderTarget | MTLTextureUsageShaderRead | MTLTextureUsageShaderWrite | MTLTextureUsagePixelFormatView; /* TODO(Metal): Optimize usage flags. */ - this->texture_descriptor_.storageMode = MTLStorageModePrivate; - this->texture_descriptor_.sampleCount = 1; - this->texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; - this->texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; + texture_descriptor_.storageMode = MTLStorageModePrivate; + texture_descriptor_.sampleCount = 1; + texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; + texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; } break; /* CUBE TEXTURES */ @@ -1771,69 +1628,63 @@ void gpu::MTLTexture::ensure_baked() case GPU_TEXTURE_CUBE_ARRAY: { /* NOTE: For a cube-map 'Texture::d_' refers to total number of faces, * not just array slices. */ - BLI_assert(this->w_ > 0 && this->h_ > 0); - this->texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; - this->texture_descriptor_.pixelFormat = mtl_format; - this->texture_descriptor_.textureType = (this->type_ == GPU_TEXTURE_CUBE_ARRAY) ? - MTLTextureTypeCubeArray : - MTLTextureTypeCube; - this->texture_descriptor_.width = this->w_; - this->texture_descriptor_.height = this->h_; - this->texture_descriptor_.depth = 1; - this->texture_descriptor_.arrayLength = (this->type_ == GPU_TEXTURE_CUBE_ARRAY) ? - this->d_ / 6 : - 1; - this->texture_descriptor_.mipmapLevelCount = (this->mtl_max_mips_ > 0) ? - this->mtl_max_mips_ : - 1; - this->texture_descriptor_.usage = + BLI_assert(w_ > 0 && h_ > 0); + texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; + texture_descriptor_.pixelFormat = mtl_format; + texture_descriptor_.textureType = (type_ == GPU_TEXTURE_CUBE_ARRAY) ? + MTLTextureTypeCubeArray : + MTLTextureTypeCube; + texture_descriptor_.width = w_; + texture_descriptor_.height = h_; + texture_descriptor_.depth = 1; + texture_descriptor_.arrayLength = (type_ == GPU_TEXTURE_CUBE_ARRAY) ? d_ / 6 : 1; + texture_descriptor_.mipmapLevelCount = (mtl_max_mips_ > 0) ? mtl_max_mips_ : 1; + texture_descriptor_.usage = MTLTextureUsageRenderTarget | MTLTextureUsageShaderRead | MTLTextureUsageShaderWrite | MTLTextureUsagePixelFormatView; /* TODO(Metal): Optimize usage flags. */ - this->texture_descriptor_.storageMode = MTLStorageModePrivate; - this->texture_descriptor_.sampleCount = 1; - this->texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; - this->texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; + texture_descriptor_.storageMode = MTLStorageModePrivate; + texture_descriptor_.sampleCount = 1; + texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; + texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; } break; /* GPU_TEXTURE_BUFFER */ case GPU_TEXTURE_BUFFER: { - this->texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; - this->texture_descriptor_.pixelFormat = mtl_format; - this->texture_descriptor_.textureType = MTLTextureTypeTextureBuffer; - this->texture_descriptor_.width = this->w_; - this->texture_descriptor_.height = 1; - this->texture_descriptor_.depth = 1; - this->texture_descriptor_.arrayLength = 1; - this->texture_descriptor_.mipmapLevelCount = (this->mtl_max_mips_ > 0) ? - this->mtl_max_mips_ : - 1; - this->texture_descriptor_.usage = + texture_descriptor_ = [[MTLTextureDescriptor alloc] init]; + texture_descriptor_.pixelFormat = mtl_format; + texture_descriptor_.textureType = MTLTextureTypeTextureBuffer; + texture_descriptor_.width = w_; + texture_descriptor_.height = 1; + texture_descriptor_.depth = 1; + texture_descriptor_.arrayLength = 1; + texture_descriptor_.mipmapLevelCount = (mtl_max_mips_ > 0) ? mtl_max_mips_ : 1; + texture_descriptor_.usage = MTLTextureUsageShaderRead | MTLTextureUsageShaderWrite | MTLTextureUsagePixelFormatView; /* TODO(Metal): Optimize usage flags. */ - this->texture_descriptor_.storageMode = MTLStorageModePrivate; - this->texture_descriptor_.sampleCount = 1; - this->texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; - this->texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; + texture_descriptor_.storageMode = MTLStorageModePrivate; + texture_descriptor_.sampleCount = 1; + texture_descriptor_.cpuCacheMode = MTLCPUCacheModeDefaultCache; + texture_descriptor_.hazardTrackingMode = MTLHazardTrackingModeDefault; } break; default: { - MTL_LOG_ERROR("[METAL] Error: Cannot create texture with unknown type: %d\n", this->type_); + MTL_LOG_ERROR("[METAL] Error: Cannot create texture with unknown type: %d\n", type_); return; } break; } /* Determine Resource Mode. */ - this->resource_mode_ = MTL_TEXTURE_MODE_DEFAULT; + resource_mode_ = MTL_TEXTURE_MODE_DEFAULT; /* Create texture. */ - this->texture_ = [ctx->device newTextureWithDescriptor:this->texture_descriptor_]; - - [this->texture_descriptor_ release]; - this->texture_descriptor_ = nullptr; - this->texture_.label = [NSString stringWithUTF8String:this->get_name()]; - BLI_assert(this->texture_); - this->is_baked_ = true; - this->is_dirty_ = false; + texture_ = [ctx->device newTextureWithDescriptor:texture_descriptor_]; + + [texture_descriptor_ release]; + texture_descriptor_ = nullptr; + texture_.label = [NSString stringWithUTF8String:this->get_name()]; + BLI_assert(texture_); + is_baked_ = true; + is_dirty_ = false; } /* Re-apply previous contents. */ @@ -1850,30 +1701,30 @@ void gpu::MTLTexture::reset() MTL_LOG_INFO("Texture %s reset. Size %d, %d, %d\n", this->get_name(), w_, h_, d_); /* Delete associated METAL resources. */ - if (this->texture_ != nil) { - [this->texture_ release]; - this->texture_ = nil; - this->is_baked_ = false; - this->is_dirty_ = true; + if (texture_ != nil) { + [texture_ release]; + texture_ = nil; + is_baked_ = false; + is_dirty_ = true; } - if (this->mip_swizzle_view_ != nil) { - [this->mip_swizzle_view_ release]; - this->mip_swizzle_view_ = nil; + if (mip_swizzle_view_ != nil) { + [mip_swizzle_view_ release]; + mip_swizzle_view_ = nil; } - if (this->texture_buffer_ != nil) { - [this->texture_buffer_ release]; + if (texture_buffer_ != nil) { + [texture_buffer_ release]; } /* Blit framebuffer. */ - if (this->blit_fb_) { - GPU_framebuffer_free(this->blit_fb_); - this->blit_fb_ = nullptr; + if (blit_fb_) { + GPU_framebuffer_free(blit_fb_); + blit_fb_ = nullptr; } - BLI_assert(this->texture_ == nil); - BLI_assert(this->mip_swizzle_view_ == nil); + BLI_assert(texture_ == nil); + BLI_assert(mip_swizzle_view_ == nil); } /** \} */ diff --git a/source/blender/gpu/metal/mtl_texture_util.mm b/source/blender/gpu/metal/mtl_texture_util.mm index 27efc770e73..e2f0b3c848e 100644 --- a/source/blender/gpu/metal/mtl_texture_util.mm +++ b/source/blender/gpu/metal/mtl_texture_util.mm @@ -493,13 +493,13 @@ void gpu::MTLTexture::update_sub_depth_2d( int mip, int offset[3], int extent[3], eGPUDataFormat type, const void *data) { /* Verify we are in a valid configuration. */ - BLI_assert(ELEM(this->format_, + BLI_assert(ELEM(format_, GPU_DEPTH_COMPONENT24, GPU_DEPTH_COMPONENT32F, GPU_DEPTH_COMPONENT16, GPU_DEPTH24_STENCIL8, GPU_DEPTH32F_STENCIL8)); - BLI_assert(validate_data_format_mtl(this->format_, type)); + BLI_assert(validate_data_format_mtl(format_, type)); BLI_assert(ELEM(type, GPU_DATA_FLOAT, GPU_DATA_UINT_24_8, GPU_DATA_UINT)); /* Determine whether we are in GPU_DATA_UINT_24_8 or GPU_DATA_FLOAT mode. */ @@ -528,7 +528,7 @@ void gpu::MTLTexture::update_sub_depth_2d( /* Push contents into an r32_tex and render contents to depth using a shader. */ GPUTexture *r32_tex_tmp = GPU_texture_create_2d( - "depth_intermediate_copy_tex", this->w_, this->h_, 1, format, nullptr); + "depth_intermediate_copy_tex", w_, h_, 1, format, nullptr); GPU_texture_filter_mode(r32_tex_tmp, false); GPU_texture_wrap_mode(r32_tex_tmp, false, true); gpu::MTLTexture *mtl_tex = static_cast<gpu::MTLTexture *>(unwrap(r32_tex_tmp)); @@ -538,7 +538,7 @@ void gpu::MTLTexture::update_sub_depth_2d( GPUFrameBuffer *depth_fb_temp = GPU_framebuffer_create("depth_intermediate_copy_fb"); GPU_framebuffer_texture_attach(depth_fb_temp, wrap(static_cast<Texture *>(this)), 0, mip); GPU_framebuffer_bind(depth_fb_temp); - if (extent[0] == this->w_ && extent[1] == this->h_) { + if (extent[0] == w_ && extent[1] == h_) { /* Skip load if the whole texture is being updated. */ GPU_framebuffer_clear_depth(depth_fb_temp, 0.0); GPU_framebuffer_clear_stencil(depth_fb_temp, 0); @@ -553,7 +553,7 @@ void gpu::MTLTexture::update_sub_depth_2d( GPU_batch_uniform_1i(quad, "mip", mip); GPU_batch_uniform_2f(quad, "extent", (float)extent[0], (float)extent[1]); GPU_batch_uniform_2f(quad, "offset", (float)offset[0], (float)offset[1]); - GPU_batch_uniform_2f(quad, "size", (float)this->w_, (float)this->h_); + GPU_batch_uniform_2f(quad, "size", (float)w_, (float)h_); bool depth_write_prev = GPU_depth_mask_get(); uint stencil_mask_prev = GPU_stencil_mask_get(); @@ -624,11 +624,11 @@ id<MTLComputePipelineState> gpu::MTLTexture::mtl_texture_read_impl( depth_scale_factor = 1; break; case 2: - /* D24 unsigned int */ + /* D24 uint */ depth_scale_factor = 0xFFFFFFu; break; case 4: - /* D32 unsigned int */ + /* D32 uint */ depth_scale_factor = 0xFFFFFFFFu; break; default: diff --git a/source/blender/gpu/opengl/gl_context.cc b/source/blender/gpu/opengl/gl_context.cc index f4f060821b0..e6af126e9cd 100644 --- a/source/blender/gpu/opengl/gl_context.cc +++ b/source/blender/gpu/opengl/gl_context.cc @@ -149,6 +149,16 @@ void GLContext::deactivate() is_active_ = false; } +void GLContext::begin_frame() +{ + /* No-op. */ +} + +void GLContext::end_frame() +{ + /* No-op. */ +} + /** \} */ /* -------------------------------------------------------------------- */ diff --git a/source/blender/gpu/opengl/gl_context.hh b/source/blender/gpu/opengl/gl_context.hh index c333c8a4afd..234bc712513 100644 --- a/source/blender/gpu/opengl/gl_context.hh +++ b/source/blender/gpu/opengl/gl_context.hh @@ -106,6 +106,8 @@ class GLContext : public Context { void activate() override; void deactivate() override; + void begin_frame() override; + void end_frame() override; void flush() override; void finish() override; diff --git a/source/blender/gpu/opengl/gl_framebuffer.hh b/source/blender/gpu/opengl/gl_framebuffer.hh index 3927ff27532..9dcdb5d13cd 100644 --- a/source/blender/gpu/opengl/gl_framebuffer.hh +++ b/source/blender/gpu/opengl/gl_framebuffer.hh @@ -77,6 +77,11 @@ class GLFrameBuffer : public FrameBuffer { eGPUDataFormat data_format, const void *clear_value) override; + /* Attachment load-stores are currently no-op's in OpenGL. */ + void attachment_set_loadstore_op(GPUAttachmentType type, + eGPULoadOp load_action, + eGPUStoreOp store_action) override{}; + void read(eGPUFrameBufferBits planes, eGPUDataFormat format, const int area[4], diff --git a/source/blender/windowmanager/intern/wm_draw.c b/source/blender/windowmanager/intern/wm_draw.c index aaa28b1fd85..b6953b21b65 100644 --- a/source/blender/windowmanager/intern/wm_draw.c +++ b/source/blender/windowmanager/intern/wm_draw.c @@ -1098,6 +1098,8 @@ static void wm_draw_window_onscreen(bContext *C, wmWindow *win, int view) static void wm_draw_window(bContext *C, wmWindow *win) { + GPU_context_begin_frame(win->gpuctx); + bScreen *screen = WM_window_get_active_screen(win); bool stereo = WM_stereo3d_enabled(win, false); @@ -1167,6 +1169,8 @@ static void wm_draw_window(bContext *C, wmWindow *win) } screen->do_draw = false; + + GPU_context_end_frame(win->gpuctx); } /** @@ -1177,8 +1181,12 @@ static void wm_draw_surface(bContext *C, wmSurface *surface) wm_window_clear_drawable(CTX_wm_manager(C)); wm_surface_make_drawable(surface); + GPU_context_begin_frame(surface->gpu_ctx); + surface->draw(C); + GPU_context_end_frame(surface->gpu_ctx); + /* Avoid interference with window drawable */ wm_surface_clear_drawable(); } diff --git a/source/blender/windowmanager/intern/wm_init_exit.c b/source/blender/windowmanager/intern/wm_init_exit.c index f77aad24719..252cfc6e143 100644 --- a/source/blender/windowmanager/intern/wm_init_exit.c +++ b/source/blender/windowmanager/intern/wm_init_exit.c @@ -310,6 +310,7 @@ void WM_init(bContext *C, int argc, const char **argv) IMB_thumb_clear_translations(); if (!G.background) { + GPU_render_begin(); #ifdef WITH_INPUT_NDOF /* Sets 3D mouse dead-zone. */ @@ -322,7 +323,10 @@ void WM_init(bContext *C, int argc, const char **argv) exit(-1); } + GPU_context_begin_frame(GPU_context_active_get()); UI_init(); + GPU_context_end_frame(GPU_context_active_get()); + GPU_render_end(); } BKE_subdiv_init(); |