Signed-off-by: Aman Karmani <a...@tmm1.net>
---
libavfilter/metal/utils.h | 35 +++++++++++++++++++
libavfilter/metal/utils.m | 73
+++++++++++++++++++++++++++++++++++++++
2 files changed, 108 insertions(+)
create mode 100644 libavfilter/metal/utils.h
create mode 100644 libavfilter/metal/utils.m
diff --git a/libavfilter/metal/utils.h b/libavfilter/metal/utils.h
new file mode 100644
index 0000000000..bd0319f63c
--- /dev/null
+++ b/libavfilter/metal/utils.h
@@ -0,0 +1,35 @@
+/*
+ * This file is part of FFmpeg.
+ *
+ * FFmpeg is free software; you can redistribute it and/or
+ * modify it under the terms of the GNU Lesser General Public
+ * License as published by the Free Software Foundation; either
+ * version 2.1 of the License, or (at your option) any later
version.
+ *
+ * FFmpeg is distributed in the hope that it will be useful,
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
GNU
+ * Lesser General Public License for more details.
+ *
+ * You should have received a copy of the GNU Lesser General Public
+ * License along with FFmpeg; if not, write to the Free Software
+ * Foundation, Inc., 51 Franklin Street, Fifth Floor, Boston, MA
02110-1301 USA
+ */
+
+#ifndef AVFILTER_METAL_UTILS_H
+#define AVFILTER_METAL_UTILS_H
+
+#include <Metal/Metal.h>
+#include <CoreVideo/CoreVideo.h>
+
+void ff_metal_compute_encoder_dispatch(id<MTLDevice> device,
+ id<MTLComputePipelineState>
pipeline,
+ id<MTLComputeCommandEncoder>
encoder,
+ NSUInteger width, NSUInteger
height);
+
+CVMetalTextureRef ff_metal_texture_from_pixbuf(void *avclass,
+
CVMetalTextureCacheRef
textureCache,
+ CVPixelBufferRef
pixbuf,
+ int plane,
+ MTLPixelFormat
format);
+#endif /* AVFILTER_METAL_UTILS_H */
diff --git a/libavfilter/metal/utils.m b/libavfilter/metal/utils.m
new file mode 100644
index 0000000000..5df0ed600e
--- /dev/null
+++ b/libavfilter/metal/utils.m
@@ -0,0 +1,73 @@
+/*
+ * This file is part of FFmpeg.
+ *
+ * FFmpeg is free software; you can redistribute it and/or
+ * modify it under the terms of the GNU Lesser General Public
+ * License as published by the Free Software Foundation; either
+ * version 2.1 of the License, or (at your option) any later
version.
+ *
+ * FFmpeg is distributed in the hope that it will be useful,
+ * but WITHOUT ANY WARRANTY; without even the implied warranty of
+ * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE. See the
GNU
+ * Lesser General Public License for more details.
+ *
+ * You should have received a copy of the GNU Lesser General Public
+ * License along with FFmpeg; if not, write to the Free Software
+ * Foundation, Inc., 51 Franklin Street, Fifth Floor, Boston, MA
02110-1301 USA
+ */
+
+#include "libavutil/log.h"
+#include <libavfilter/metal/utils.h>
+
+void ff_metal_compute_encoder_dispatch(id<MTLDevice> device,
+ id<MTLComputePipelineState>
pipeline,
+ id<MTLComputeCommandEncoder>
encoder,
+ NSUInteger width, NSUInteger
height)
+{
+ [encoder setComputePipelineState:pipeline];
+ NSUInteger w = pipeline.threadExecutionWidth;
+ NSUInteger h = pipeline.maxTotalThreadsPerThreadgroup / w;
+ MTLSize threadsPerThreadgroup = MTLSizeMake(w, h, 1);
+ BOOL fallback = YES;
+ if (@available(macOS 10.15, iOS 11, tvOS 14.5, *)) {
+ if ([device supportsFamily:MTLGPUFamilyCommon3]) {
+ MTLSize threadsPerGrid = MTLSizeMake(width, height, 1);
+ [encoder dispatchThreads:threadsPerGrid
threadsPerThreadgroup:threadsPerThreadgroup];
+ fallback = NO;
+ }
+ }
