From 802697fdecfd767f87deaa4e29d31fcbc0b66a7c Mon Sep 17 00:00:00 2001 From: Stuart Carnie Date: Tue, 17 Jul 2018 20:59:06 -0700 Subject: [PATCH] fix(Metal): Use textures for pixel conversion --- gfx/common/metal/Context.h | 2 +- gfx/common/metal/Context.m | 34 ++++++++++---- gfx/common/metal/Shaders.metal | 54 ++++++++++----------- gfx/common/metal/TexturedView.m | 42 ++++++----------- gfx/common/metal_common.m | 83 +++++++++++++++------------------ ui/drivers/ui_cocoa.m | 2 + 6 files changed, 102 insertions(+), 115 deletions(-) diff --git a/gfx/common/metal/Context.h b/gfx/common/metal/Context.h index 497da8e44d..62a799a389 100644 --- a/gfx/common/metal/Context.h +++ b/gfx/common/metal/Context.h @@ -52,7 +52,7 @@ typedef struct - (Texture *)newTexture:(struct texture_image)image filter:(enum texture_filter_type)filter; - (id)newTexture:(struct texture_image)image mipmapped:(bool)mipmapped; -- (void)convertFormat:(RPixelFormat)fmt from:(id)src to:(id)dst; +- (void)convertFormat:(RPixelFormat)fmt from:(id)src to:(id)dst; - (id)getStockShader:(int)index blend:(bool)blend; /*! @brief resets the viewport for the main render encoder to the drawable size */ diff --git a/gfx/common/metal/Context.m b/gfx/common/metal/Context.m index f610320c24..70f7f66643 100644 --- a/gfx/common/metal/Context.m +++ b/gfx/common/metal/Context.m @@ -65,7 +65,9 @@ _inflightSemaphore = dispatch_semaphore_create(MAX_INFLIGHT); _device = d; _layer = layer; +#if TARGET_OS_OSX _layer.displaySyncEnabled = YES; +#endif _library = l; _commandQueue = [_device newCommandQueue]; _clearColor = MTLClearColorMake(0, 0, 0, 1); @@ -127,7 +129,16 @@ - (void)setDisplaySyncEnabled:(bool)displaySyncEnabled { +#if TARGET_OS_OSX _layer.displaySyncEnabled = displaySyncEnabled; +#endif +} + +- (bool)displaySyncEnabled +{ +#if TARGET_OS_OSX + return _layer.displaySyncEnabled; +#endif } #pragma mark - shaders @@ -154,11 +165,6 @@ return _states[index][blend ? 1 : 0]; } -- (bool)displaySyncEnabled -{ - return _layer.displaySyncEnabled; -} - - (MTLVertexDescriptor *)_spriteVertexDescriptor { MTLVertexDescriptor *vd = [MTLVertexDescriptor new]; @@ -441,13 +447,13 @@ return _drawable; } -- (void)convertFormat:(RPixelFormat)fmt from:(id)src to:(id)dst +- (void)convertFormat:(RPixelFormat)fmt from:(id)src to:(id)dst { - assert(dst.width * dst.height == src.length / RPixelFormatToBPP(fmt)); + assert(src.width == dst.width && src.height == dst.height); assert(fmt >= 0 && fmt < RPixelFormatCount); Filter *conv = _filters[fmt]; assert(conv != nil); - [conv apply:self.blitCommandBuffer inBuf:src outTex:dst]; + [conv apply:self.blitCommandBuffer in:src out:dst]; } - (id)blitCommandBuffer @@ -615,6 +621,7 @@ static const NSUInteger kConstantAlignment = 4; - (void)commitRanges { +#if TARGET_OS_OSX for (BufferNode *n = _head; n != nil; n = n.next) { if (n.allocated > 0) @@ -622,6 +629,7 @@ static const NSUInteger kConstantAlignment = 4; [n.src didModifyRange:NSMakeRange(0, n.allocated)]; } } +#endif } - (void)discard @@ -635,9 +643,15 @@ static const NSUInteger kConstantAlignment = 4; { bzero(range, sizeof(*range)); +#if TARGET_OS_OSX + MTLResourceOptions opts = MTLResourceStorageModeManaged; +#else + MTLResourceOptions opts = MTLResourceStorageModeShared; +#endif + if (!_head) { - _head = [[BufferNode alloc] initWithBuffer:[_device newBufferWithLength:_blockLen options:MTLResourceStorageModeManaged]]; + _head = [[BufferNode alloc] initWithBuffer:[_device newBufferWithLength:_blockLen options:opts]]; _length += _blockLen; _current = _head; _offset = 0; @@ -659,7 +673,7 @@ static const NSUInteger kConstantAlignment = 4; blockLen = length; } - _current.next = [[BufferNode alloc] initWithBuffer:[_device newBufferWithLength:blockLen options:MTLResourceStorageModeManaged]]; + _current.next = [[BufferNode alloc] initWithBuffer:[_device newBufferWithLength:blockLen options:opts]]; if (!_current.next) return NO; diff --git a/gfx/common/metal/Shaders.metal b/gfx/common/metal/Shaders.metal index c5449b2989..396eafd050 100644 --- a/gfx/common/metal/Shaders.metal +++ b/gfx/common/metal/Shaders.metal @@ -81,38 +81,32 @@ fragment half4 stock_fragment_color(FontFragmentIn in [[ stage_in ]]) #pragma mark - filter kernels -kernel void convert_bgra4444_to_bgra8888(device uint16_t * in [[ buffer(0) ]], - texture2d out [[ texture(0) ]], - uint id [[ thread_position_in_grid ]]) +kernel void convert_bgra4444_to_bgra8888(texture2d in [[ texture(0) ]], + texture2d out [[ texture(1) ]], + uint2 gid [[ thread_position_in_grid ]]) { - uint16_t pix = in[id]; - uchar4 pix2 = uchar4( - extract_bits(pix, 4, 4), - extract_bits(pix, 8, 4), - extract_bits(pix, 12, 4), - extract_bits(pix, 0, 4) - ); - - uint ypos = id / out.get_width(); - uint xpos = id % out.get_width(); - - out.write(half4(pix2) / 15.0, uint2(xpos, ypos)); + ushort pix = in.read(gid).r; + uchar4 pix2 = uchar4( + extract_bits(pix, 4, 4), + extract_bits(pix, 8, 4), + extract_bits(pix, 12, 4), + extract_bits(pix, 0, 4) + ); + + out.write(half4(pix2) / 15.0, gid); } -kernel void convert_rgb565_to_bgra8888(device uint16_t * in [[ buffer(0) ]], - texture2d out [[ texture(0) ]], - uint id [[ thread_position_in_grid ]]) +kernel void convert_rgb565_to_bgra8888(texture2d in [[ texture(0) ]], + texture2d out [[ texture(1) ]], + uint2 gid [[ thread_position_in_grid ]]) { - uint16_t pix = in[id]; - uchar4 pix2 = uchar4( - extract_bits(pix, 11, 5), - extract_bits(pix, 5, 6), - extract_bits(pix, 0, 5), - 0xf - ); - - uint ypos = id / out.get_width(); - uint xpos = id % out.get_width(); - - out.write(half4(pix2) / half4(0x1f, 0x3f, 0x1f, 0xf), uint2(xpos, ypos)); + ushort pix = in.read(gid).r; + uchar4 pix2 = uchar4( + extract_bits(pix, 11, 5), + extract_bits(pix, 5, 6), + extract_bits(pix, 0, 5), + 0xf + ); + + out.write(half4(pix2) / half4(0x1f, 0x3f, 0x1f, 0xf), gid); } diff --git a/gfx/common/metal/TexturedView.m b/gfx/common/metal/TexturedView.m index f89346a654..89859ec2b0 100644 --- a/gfx/common/metal/TexturedView.m +++ b/gfx/common/metal/TexturedView.m @@ -16,8 +16,8 @@ CGRect _frame; NSUInteger _bpp; - id _pixels; // frame buffer in _srcFmt - bool _pixelsDirty; + id _src; // source texture + bool _srcDirty; } - (instancetype)initWithDescriptor:(ViewDescriptor *)d context:(Context *)c @@ -53,7 +53,6 @@ _size = size; - // create new texture { MTLTextureDescriptor *td = [MTLTextureDescriptor texture2DDescriptorWithPixelFormat:MTLPixelFormatBGRA8Unorm width:(NSUInteger)size.width @@ -65,8 +64,11 @@ if (_format != RPixelFormatBGRA8Unorm && _format != RPixelFormatBGRX8Unorm) { - _pixels = [_context.device newBufferWithLength:(NSUInteger)(size.width * size.height * 2) - options:MTLResourceStorageModeManaged]; + MTLTextureDescriptor *td = [MTLTextureDescriptor texture2DDescriptorWithPixelFormat:MTLPixelFormatR16Uint + width:(NSUInteger)size.width + height:(NSUInteger)size.height + mipmapped:NO]; + _src = [_context.device newTextureWithDescriptor:td]; } } @@ -112,11 +114,11 @@ if (_format == RPixelFormatBGRA8Unorm || _format == RPixelFormatBGRX8Unorm) return; - if (!_pixelsDirty) + if (!_srcDirty) return; - [_context convertFormat:_format from:_pixels to:_texture]; - _pixelsDirty = NO; + [_context convertFormat:_format from:_src to:_texture]; + _srcDirty = NO; } - (void)drawWithContext:(Context *)ctx @@ -141,26 +143,10 @@ } else { - void *dst = _pixels.contents; - size_t len = (size_t)(_bpp * _size.width); - assert(len <= pitch); // the length can't be larger? - - if (len < pitch) - { - for (int i = 0; i < _size.height; i++) - { - memcpy(dst, src, len); - dst += len; - src += pitch; - } - } - else - { - memcpy(dst, src, _pixels.length); - } - - [_pixels didModifyRange:NSMakeRange(0, _pixels.length)]; - _pixelsDirty = YES; + [_src replaceRegion:MTLRegionMake2D(0, 0, (NSUInteger)_size.width, (NSUInteger)_size.height) + mipmapLevel:0 withBytes:src + bytesPerRow:(NSUInteger)(pitch)]; + _srcDirty = YES; } } diff --git a/gfx/common/metal_common.m b/gfx/common/metal_common.m index 007f451abf..f1c146bd40 100644 --- a/gfx/common/metal_common.m +++ b/gfx/common/metal_common.m @@ -299,30 +299,29 @@ settings_t *settings = config_get_ptr(); if (settings && settings->bools.video_msg_bgcolor_enable) { - int msg_width = + int msg_width = font_driver_get_message_width(NULL, msg, (unsigned)strlen(msg), 1.0f); - - float x = video_info->font_msg_pos_x; - float y = 1.0f - video_info->font_msg_pos_y; - float width = msg_width / (float)_viewport->full_width; - float height = + + float x = video_info->font_msg_pos_x; + float y = 1.0f - video_info->font_msg_pos_y; + float width = msg_width / (float)_viewport->full_width; + float height = settings->floats.video_font_size / (float)_viewport->full_height; y -= height; - - float x2 = 0.005f; /* extend background around text */ - float y2 = 0.005f; - - x -= x2; - y -= y2; - width += x2; - height += y2; - - float r = settings->uints.video_msg_bgcolor_red / 255.0f; - float g = settings->uints.video_msg_bgcolor_green / 255.0f; - float b = settings->uints.video_msg_bgcolor_blue / 255.0f; - float a = settings->floats.video_msg_bgcolor_opacity; + float x2 = 0.005f; /* extend background around text */ + float y2 = 0.005f; + + x -= x2; + y -= y2; + width += x2; + height += y2; + + float r = settings->uints.video_msg_bgcolor_red / 255.0f; + float g = settings->uints.video_msg_bgcolor_green / 255.0f; + float b = settings->uints.video_msg_bgcolor_blue / 255.0f; + float a = settings->floats.video_msg_bgcolor_opacity; [_context resetRenderViewport]; [_context drawQuadX:x y:y w:width h:height r:r g:g b:b a:a]; } @@ -332,7 +331,12 @@ - (void)_beginFrame { + video_viewport_t vp = *_viewport; video_driver_update_viewport(_viewport, NO, _keepAspect); + if (memcmp(&vp, _viewport, sizeof(vp)) != 0) + { + _context.viewport = _viewport; + } [_context begin]; [self _updateUniforms]; } @@ -546,8 +550,8 @@ typedef struct MTLALIGN(16) CGRect _frame; NSUInteger _bpp; - id _pixels; // frame buffer in _srcFmt - bool _pixelsDirty; + id _src; // src texture + bool _srcDirty; id _samplers[RARCH_FILTER_MAX][RARCH_WRAP_MAX]; struct video_shader *_shader; @@ -655,8 +659,11 @@ typedef struct MTLALIGN(16) if (_format != RPixelFormatBGRA8Unorm && _format != RPixelFormatBGRX8Unorm) { - _pixels = [_context.device newBufferWithLength:(NSUInteger)(size.width * size.height * 2) - options:MTLResourceStorageModeManaged]; + MTLTextureDescriptor *td = [MTLTextureDescriptor texture2DDescriptorWithPixelFormat:MTLPixelFormatR16Uint + width:(NSUInteger)size.width + height:(NSUInteger)size.height + mipmapped:NO]; + _src = [_context.device newTextureWithDescriptor:td]; } } @@ -702,11 +709,11 @@ typedef struct MTLALIGN(16) if (_format == RPixelFormatBGRA8Unorm || _format == RPixelFormatBGRX8Unorm) return; - if (!_pixelsDirty) + if (!_srcDirty) return; - [_context convertFormat:_format from:_pixels to:_texture]; - _pixelsDirty = NO; + [_context convertFormat:_format from:_src to:_texture]; + _srcDirty = NO; } - (void)_updateHistory @@ -778,26 +785,10 @@ typedef struct MTLALIGN(16) } else { - void *dst = _pixels.contents; - size_t len = (size_t)(_bpp * _size.width); - assert(len <= pitch); // the length can't be larger? - - if (len < pitch) - { - for (int i = 0; i < _size.height; i++) - { - memcpy(dst, src, len); - dst += len; - src += pitch; - } - } - else - { - memcpy(dst, src, _pixels.length); - } - - [_pixels didModifyRange:NSMakeRange(0, _pixels.length)]; - _pixelsDirty = YES; + [_src replaceRegion:MTLRegionMake2D(0, 0, (NSUInteger)_size.width, (NSUInteger)_size.height) + mipmapLevel:0 withBytes:src + bytesPerRow:(NSUInteger)(pitch)]; + _srcDirty = YES; } } diff --git a/ui/drivers/ui_cocoa.m b/ui/drivers/ui_cocoa.m index bf66477fe6..5de129189a 100644 --- a/ui/drivers/ui_cocoa.m +++ b/ui/drivers/ui_cocoa.m @@ -297,6 +297,8 @@ static char** waiting_argv; [self performSelectorOnMainThread:@selector(rarch_main) withObject:nil waitUntilDone:NO]; } +#pragma mark - ApplePlatform + - (void)setViewType:(apple_view_type_t)vt { if (vt == _vt) { return;