FaSTTUBe/GST-Tensordecoder-ov_ep
37
0
Fork 0
Code Issues Pull Requests Actions Packages Projects Releases Wiki Activity

hip: Add AMD HIP plugin

Browse Source 
Adding hipupload, hipdownload, and hipconvert family elements

Part-of: <https://gitlab.freedesktop.org/gstreamer/gstreamer/-/merge_requests/8923>
This commit is contained in:
Seungha Yang 2025-04-26 03:24:37 +09:00
parent 79677043dd
commit 0f9ce9de90
27 changed files with 10227 additions and 0 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

1
.indent_cpp_list
View File

@ -12,6 +12,7 @@ subprojects/gst-plugins-bad/sys/amfcodec
subprojects/gst-plugins-bad/sys/d3d11
subprojects/gst-plugins-bad/sys/d3d12
subprojects/gst-plugins-bad/sys/dwrite
subprojects/gst-plugins-bad/sys/hip
subprojects/gst-plugins-bad/sys/mediafoundation
subprojects/gst-plugins-bad/sys/nvcodec
^(subprojects/gst-plugins-bad/sys/qsv/)+(\w)+([^/])+(cpp$)

1
subprojects/gst-plugins-bad/meson.options
View File

@ -127,6 +127,7 @@ option('gme', type : 'feature', value : 'auto', description : 'libgme gaming con
option('gs', type : 'feature', value : 'auto', description : 'Google Cloud Storage source and sink plugin')
option('gsm', type : 'feature', value : 'auto', description : 'GSM encoder/decoder plugin')
option('gtk3', type : 'feature', value : 'auto', description : 'GTK+ video sink plugin')
option('hip', type : 'feature', value : 'auto', description : 'AMD HIP plugin')
option('ipcpipeline', type : 'feature', value : 'auto', description : 'Inter-process communication plugin')
option('iqa', type : 'feature', value : 'auto', description : 'Image quality assessment plugin (AGPL - only built if gpl option is also enabled!)')
option('kms', type : 'feature', value : 'auto', description : 'KMS video sink plugin')

30
subprojects/gst-plugins-bad/sys/hip/gsthip.h Normal file
View File

@ -0,0 +1,30 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include <hip/hip_runtime.h>
#include "gsthip_fwd.h"
#include "gsthipdevice.h"
#include "gsthipmemory.h"
#include "gsthipbufferpool.h"
#include "gsthiputils.h"

46
subprojects/gst-plugins-bad/sys/hip/gsthip_fwd.h Normal file
View File

@ -0,0 +1,46 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
G_BEGIN_DECLS
typedef struct _GstHipDevice GstHipDevice;
typedef struct _GstHipDeviceClass GstHipDeviceClass;
typedef struct _GstHipDevicePrivate GstHipDevicePrivate;
typedef struct _GstHipMemory GstHipMemory;
typedef struct _GstHipMemoryPrivate GstHipMemoryPrivate;
typedef struct _GstHipAllocator GstHipAllocator;
typedef struct _GstHipAllocatorClass GstHipAllocatorClass;
typedef struct _GstHipAllocatorPrivate GstHipAllocatorPrivate;
typedef struct _GstHipPoolAllocator GstHipPoolAllocator;
typedef struct _GstHipPoolAllocatorClass GstHipPoolAllocatorClass;
typedef struct _GstHipPoolAllocatorPrivate GstHipPoolAllocatorPrivate;
typedef struct _GstHipBufferPool GstHipBufferPool;
typedef struct _GstHipBufferPoolClass GstHipBufferPoolClass;
typedef struct _GstHipBufferPoolPrivate GstHipBufferPoolPrivate;
G_END_DECLS

349
subprojects/gst-plugins-bad/sys/hip/gsthipbasefilter.cpp Normal file
View File

@ -0,0 +1,349 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include "config.h"
#endif
#include "gsthipbasefilter.h"
#include <mutex>
GST_DEBUG_CATEGORY_STATIC (gst_hip_base_filter_debug);
#define GST_CAT_DEFAULT gst_hip_base_filter_debug
/* cached quark to avoid contention on the global quark table lock */
#define META_TAG_VIDEO meta_tag_video_quark
static GQuark meta_tag_video_quark;
enum
{
PROP_0,
PROP_DEVICE_ID,
};
#define DEFAULT_DEVICE_ID -1
struct _GstHipBaseFilterPrivate
{
~_GstHipBaseFilterPrivate ()
{
gst_clear_caps (&in_caps);
gst_clear_caps (&out_caps);
}
std::recursive_mutex lock;
GstCaps *in_caps = nullptr;
GstCaps *out_caps = nullptr;
gint device_id = DEFAULT_DEVICE_ID;
};
#define gst_hip_base_filter_parent_class parent_class
G_DEFINE_ABSTRACT_TYPE (GstHipBaseFilter, gst_hip_base_filter,
GST_TYPE_BASE_TRANSFORM);
static void gst_hip_base_filter_set_property (GObject * object,
guint prop_id, const GValue * value, GParamSpec * pspec);
static void gst_hip_base_filter_get_property (GObject * object,
guint prop_id, GValue * value, GParamSpec * pspec);
static void gst_hip_base_filter_finalize (GObject * object);
static void gst_hip_base_filter_set_context (GstElement * element,
GstContext * context);
static gboolean gst_hip_base_filter_start (GstBaseTransform * trans);
static gboolean gst_hip_base_filter_stop (GstBaseTransform * trans);
static gboolean gst_hip_base_filter_set_caps (GstBaseTransform * trans,
GstCaps * incaps, GstCaps * outcaps);
static gboolean gst_hip_base_filter_get_unit_size (GstBaseTransform * trans,
GstCaps * caps, gsize * size);
static gboolean gst_hip_base_filter_query (GstBaseTransform * trans,
GstPadDirection direction, GstQuery * query);
static void gst_hip_base_filter_before_transform (GstBaseTransform * trans,
GstBuffer * buffer);
static gboolean
gst_hip_base_filter_transform_meta (GstBaseTransform * trans,
GstBuffer * outbuf, GstMeta * meta, GstBuffer * inbuf);
static void
gst_hip_base_filter_class_init (GstHipBaseFilterClass * klass)
{
auto object_class = G_OBJECT_CLASS (klass);
auto element_class = GST_ELEMENT_CLASS (klass);
auto trans_class = GST_BASE_TRANSFORM_CLASS (klass);
object_class->finalize = gst_hip_base_filter_finalize;
object_class->set_property = gst_hip_base_filter_set_property;
object_class->get_property = gst_hip_base_filter_get_property;
g_object_class_install_property (object_class, PROP_DEVICE_ID,
g_param_spec_int ("device-id",
"Device ID", "HIP device ID to use (-1 = auto)",
-1, G_MAXINT, DEFAULT_DEVICE_ID,
(GParamFlags) (G_PARAM_READWRITE | GST_PARAM_MUTABLE_READY |
G_PARAM_STATIC_STRINGS)));
element_class->set_context =
GST_DEBUG_FUNCPTR (gst_hip_base_filter_set_context);
trans_class->passthrough_on_same_caps = TRUE;
trans_class->start = GST_DEBUG_FUNCPTR (gst_hip_base_filter_start);
trans_class->stop = GST_DEBUG_FUNCPTR (gst_hip_base_filter_stop);
trans_class->set_caps = GST_DEBUG_FUNCPTR (gst_hip_base_filter_set_caps);
trans_class->get_unit_size =
GST_DEBUG_FUNCPTR (gst_hip_base_filter_get_unit_size);
trans_class->query = GST_DEBUG_FUNCPTR (gst_hip_base_filter_query);
trans_class->before_transform =
GST_DEBUG_FUNCPTR (gst_hip_base_filter_before_transform);
trans_class->transform_meta =
GST_DEBUG_FUNCPTR (gst_hip_base_filter_transform_meta);
GST_DEBUG_CATEGORY_INIT (gst_hip_base_filter_debug,
"hipbasefilter", 0, "hipbasefilter");
gst_type_mark_as_plugin_api (GST_TYPE_HIP_BASE_FILTER, (GstPluginAPIFlags) 0);
meta_tag_video_quark = g_quark_from_static_string (GST_META_TAG_VIDEO_STR);
}
static void
gst_hip_base_filter_init (GstHipBaseFilter * self)
{
self->priv = new GstHipBaseFilterPrivate ();
}
static void
gst_hip_base_filter_finalize (GObject * object)
{
auto self = GST_HIP_BASE_FILTER (object);
gst_clear_object (&self->device);
delete self->priv;
G_OBJECT_CLASS (parent_class)->finalize (object);
}
static void
gst_hip_base_filter_set_property (GObject * object, guint prop_id,
const GValue * value, GParamSpec * pspec)
{
auto self = GST_HIP_BASE_FILTER (object);
auto priv = self->priv;
std::lock_guard < std::recursive_mutex > lk (priv->lock);
switch (prop_id) {
case PROP_DEVICE_ID:
priv->device_id = g_value_get_int (value);
break;
default:
G_OBJECT_WARN_INVALID_PROPERTY_ID (object, prop_id, pspec);
break;
}
}
static void
gst_hip_base_filter_get_property (GObject * object, guint prop_id,
GValue * value, GParamSpec * pspec)
{
auto self = GST_HIP_BASE_FILTER (object);
auto priv = self->priv;
std::lock_guard < std::recursive_mutex > lk (priv->lock);
switch (prop_id) {
case PROP_DEVICE_ID:
g_value_set_int (value, priv->device_id);
break;
default:
G_OBJECT_WARN_INVALID_PROPERTY_ID (object, prop_id, pspec);
break;
}
}
static void
gst_hip_base_filter_set_context (GstElement * element, GstContext * context)
{
auto self = GST_HIP_BASE_FILTER (element);
auto priv = self->priv;
{
std::lock_guard < std::recursive_mutex > lk (priv->lock);
gst_hip_handle_set_context (element, context, priv->device_id,
&self->device);
}
GST_ELEMENT_CLASS (parent_class)->set_context (element, context);
}
static gboolean
gst_hip_base_filter_start (GstBaseTransform * trans)
{
auto self = GST_HIP_BASE_FILTER (trans);
auto priv = self->priv;
{
std::lock_guard < std::recursive_mutex > lk (priv->lock);
if (!gst_hip_ensure_element_data (GST_ELEMENT (trans), priv->device_id,
&self->device)) {
GST_ERROR_OBJECT (self, "Couldn't get HIP device");
return FALSE;
}
}
return TRUE;
}
static gboolean
gst_hip_base_filter_stop (GstBaseTransform * trans)
{
auto self = GST_HIP_BASE_FILTER (trans);
auto priv = self->priv;
{
std::lock_guard < std::recursive_mutex > lk (priv->lock);
gst_clear_object (&self->device);
gst_clear_caps (&priv->in_caps);
gst_clear_caps (&priv->out_caps);
}
return TRUE;
}
static gboolean
gst_hip_base_filter_set_caps (GstBaseTransform * trans, GstCaps * incaps,
GstCaps * outcaps)
{
auto self = GST_HIP_BASE_FILTER (trans);
auto priv = self->priv;
GstVideoInfo in_info, out_info;
if (!self->device) {
GST_ERROR_OBJECT (self, "HIP device is not configured");
return FALSE;
}
/* input caps */
if (!gst_video_info_from_caps (&in_info, incaps)) {
GST_ERROR_OBJECT (self, "invalid incaps %" GST_PTR_FORMAT, incaps);
return FALSE;
}
/* output caps */
if (!gst_video_info_from_caps (&out_info, outcaps)) {
GST_ERROR_OBJECT (self, "invalid incaps %" GST_PTR_FORMAT, incaps);
return FALSE;
}
self->in_info = in_info;
self->out_info = out_info;
gst_caps_replace (&priv->in_caps, incaps);
gst_caps_replace (&priv->out_caps, outcaps);
auto klass = GST_HIP_BASE_FILTER_GET_CLASS (self);
if (klass->set_info)
return klass->set_info (self, incaps, &in_info, outcaps, &out_info);
return TRUE;
}
static gboolean
gst_hip_base_filter_get_unit_size (GstBaseTransform * trans, GstCaps * caps,
gsize * size)
{
GstVideoInfo info;
if (!gst_video_info_from_caps (&info, caps))
return FALSE;
*size = GST_VIDEO_INFO_SIZE (&info);
return TRUE;
}
static gboolean
gst_hip_base_filter_query (GstBaseTransform * trans,
GstPadDirection direction, GstQuery * query)
{
auto self = GST_HIP_BASE_FILTER (trans);
auto priv = self->priv;
switch (GST_QUERY_TYPE (query)) {
case GST_QUERY_CONTEXT:
{
std::lock_guard < std::recursive_mutex > lk (priv->lock);
if (gst_hip_handle_context_query (GST_ELEMENT (self), query,
self->device)) {
return TRUE;
}
break;
}
default:
break;
}
return GST_BASE_TRANSFORM_CLASS (parent_class)->query (trans, direction,
query);
}
static void
gst_hip_base_filter_before_transform (GstBaseTransform * trans,
GstBuffer * buffer)
{
auto self = GST_HIP_BASE_FILTER (trans);
auto priv = self->priv;
auto mem = gst_buffer_peek_memory (buffer, 0);
if (!gst_is_hip_memory (mem))
return;
auto hmem = GST_HIP_MEMORY_CAST (mem);
/* Same context, nothing to do */
if (gst_hip_device_is_equal (self->device, hmem->device))
return;
GST_INFO_OBJECT (self, "Updating device %" GST_PTR_FORMAT " -> %"
GST_PTR_FORMAT, self->device, hmem->device);
{
std::lock_guard < std::recursive_mutex > lk (priv->lock);
gst_clear_object (&self->device);
self->device = (GstHipDevice *) gst_object_ref (hmem->device);
}
/* subclass will update internal object.
* Note that gst_base_transform_reconfigure() might not trigger this
* unless caps was changed meanwhile */
gst_hip_base_filter_set_caps (trans, priv->in_caps, priv->out_caps);
/* Mark reconfigure so that we can update pool */
gst_base_transform_reconfigure_src (trans);
}
static gboolean
gst_hip_base_filter_transform_meta (GstBaseTransform * trans,
GstBuffer * outbuf, GstMeta * meta, GstBuffer * inbuf)
{
auto info = meta->info;
auto tags = gst_meta_api_type_get_tags (info->api);
if (!tags || (g_strv_length ((gchar **) tags) == 1
&& gst_meta_api_type_has_tag (info->api, META_TAG_VIDEO)))
return TRUE;
return GST_BASE_TRANSFORM_CLASS (parent_class)->transform_meta (trans, outbuf,
meta, inbuf);
}

68
subprojects/gst-plugins-bad/sys/hip/gsthipbasefilter.h Normal file
View File

@ -0,0 +1,68 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include <gst/base/gstbasetransform.h>
#include <gst/video/video.h>
#include "gsthip.h"
G_BEGIN_DECLS
#define GST_TYPE_HIP_BASE_FILTER (gst_hip_base_filter_get_type())
#define GST_HIP_BASE_FILTER(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj),GST_TYPE_HIP_BASE_FILTER,GstHipBaseFilter))
#define GST_HIP_BASE_FILTER_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass), GST_TYPE_HIP_BASE_FILTER,GstHipBaseFilterClass))
#define GST_HIP_BASE_FILTER_GET_CLASS(obj) (G_TYPE_INSTANCE_GET_CLASS((obj), GST_TYPE_HIP_BASE_FILTER,GstHipBaseFilterClass))
#define GST_IS_HIP_BASE_FILTER(obj) (G_TYPE_CHECK_INSTANCE_TYPE((obj),GST_TYPE_HIP_BASE_FILTER))
#define GST_IS_HIP_BASE_FILTER_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE((klass), GST_TYPE_HIP_BASE_FILTER))
typedef struct _GstHipBaseFilter GstHipBaseFilter;
typedef struct _GstHipBaseFilterClass GstHipBaseFilterClass;
typedef struct _GstHipBaseFilterPrivate GstHipBaseFilterPrivate;
struct _GstHipBaseFilter
{
GstBaseTransform parent;
GstHipDevice *device;
GstVideoInfo in_info;
GstVideoInfo out_info;
GstHipBaseFilterPrivate *priv;
};
struct _GstHipBaseFilterClass
{
GstBaseTransformClass parent_class;
gboolean (*set_info) (GstHipBaseFilter *filter,
GstCaps *incaps,
GstVideoInfo *in_info,
GstCaps *outcaps,
GstVideoInfo *out_info);
};
GType gst_hip_base_filter_get_type (void);
G_DEFINE_AUTOPTR_CLEANUP_FUNC(GstHipBaseFilter, gst_object_unref)
G_END_DECLS

228
subprojects/gst-plugins-bad/sys/hip/gsthipbufferpool.cpp Normal file
View File

@ -0,0 +1,228 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include "config.h"
#endif
#include "gsthip.h"
GST_DEBUG_CATEGORY_STATIC (gst_hip_buffer_pool_debug);
#define GST_CAT_DEFAULT gst_hip_buffer_pool_debug
struct _GstHipBufferPoolPrivate
{
GstVideoInfo info;
GstHipPoolAllocator *alloc = nullptr;
};
#define gst_hip_buffer_pool_parent_class parent_class
G_DEFINE_TYPE (GstHipBufferPool, gst_hip_buffer_pool, GST_TYPE_BUFFER_POOL);
static void gst_hip_buffer_pool_finalize (GObject * object);
static const gchar **gst_hip_buffer_pool_get_options (GstBufferPool * pool);
static gboolean gst_hip_buffer_pool_set_config (GstBufferPool * pool,
GstStructure * config);
static gboolean gst_hip_buffer_pool_start (GstBufferPool * pool);
static gboolean gst_hip_buffer_pool_stop (GstBufferPool * pool);
static GstFlowReturn gst_hip_buffer_pool_alloc (GstBufferPool * pool,
GstBuffer ** buffer, GstBufferPoolAcquireParams * params);
static void
gst_hip_buffer_pool_class_init (GstHipBufferPoolClass * klass)
{
auto object_class = G_OBJECT_CLASS (klass);
auto pool_class = GST_BUFFER_POOL_CLASS (klass);
object_class->finalize = gst_hip_buffer_pool_finalize;
pool_class->get_options = gst_hip_buffer_pool_get_options;
pool_class->set_config = gst_hip_buffer_pool_set_config;
pool_class->start = gst_hip_buffer_pool_start;
pool_class->stop = gst_hip_buffer_pool_stop;
pool_class->alloc_buffer = gst_hip_buffer_pool_alloc;
GST_DEBUG_CATEGORY_INIT (gst_hip_buffer_pool_debug, "hipbufferpool", 0,
"hipbufferpool");
}
static void
gst_hip_buffer_pool_init (GstHipBufferPool * self)
{
self->priv = new GstHipBufferPoolPrivate ();
}
static void
gst_hip_buffer_pool_finalize (GObject * object)
{
auto self = GST_HIP_BUFFER_POOL (object);
auto priv = self->priv;
if (priv->alloc) {
gst_hip_allocator_set_active (GST_HIP_ALLOCATOR (priv->alloc), FALSE);
gst_clear_object (&priv->alloc);
}
gst_clear_object (&self->device);
delete self->priv;
G_OBJECT_CLASS (parent_class)->finalize (object);
}
static const gchar **
gst_hip_buffer_pool_get_options (GstBufferPool * pool)
{
static const gchar *options[] = { GST_BUFFER_POOL_OPTION_VIDEO_META, nullptr
};
return options;
}
static gboolean
gst_hip_buffer_pool_set_config (GstBufferPool * pool, GstStructure * config)
{
auto self = GST_HIP_BUFFER_POOL (pool);
auto priv = self->priv;
GstCaps *caps = nullptr;
guint size, min_buffers, max_buffers;
GstVideoInfo info;
GstMemory *mem = nullptr;
if (!gst_buffer_pool_config_get_params (config, &caps, &size, &min_buffers,
&max_buffers)) {
GST_WARNING_OBJECT (self, "invalid config");
return FALSE;
}
if (!caps) {
GST_WARNING_OBJECT (pool, "no caps in config");
return FALSE;
}
if (!gst_video_info_from_caps (&info, caps)) {
GST_WARNING_OBJECT (self, "Failed to convert caps to video-info");
return FALSE;
}
if (priv->alloc) {
gst_hip_allocator_set_active (GST_HIP_ALLOCATOR (priv->alloc), FALSE);
gst_clear_object (&priv->alloc);
}
priv->alloc = gst_hip_pool_allocator_new (self->device, &info, nullptr);
if (!priv->alloc) {
GST_ERROR_OBJECT (self, "Couldn't create allocator");
return FALSE;
}
if (!gst_hip_allocator_set_active (GST_HIP_ALLOCATOR (priv->alloc), TRUE)) {
GST_ERROR_OBJECT (self, "Couldn't set active");
return FALSE;
}
gst_hip_pool_allocator_acquire_memory (priv->alloc, &mem);
gst_hip_allocator_set_active (GST_HIP_ALLOCATOR (priv->alloc), FALSE);
if (!mem) {
GST_WARNING_OBJECT (self, "Failed to allocate memory");
return FALSE;
}
auto hmem = GST_HIP_MEMORY_CAST (mem);
gst_buffer_pool_config_set_params (config, caps,
GST_VIDEO_INFO_SIZE (&hmem->info), min_buffers, max_buffers);
priv->info = info;
gst_memory_unref (mem);
return GST_BUFFER_POOL_CLASS (parent_class)->set_config (pool, config);
}
static GstFlowReturn
gst_hip_buffer_pool_alloc (GstBufferPool * pool, GstBuffer ** buffer,
GstBufferPoolAcquireParams * params)
{
auto self = GST_HIP_BUFFER_POOL (pool);
auto priv = self->priv;
GstVideoInfo *info = &priv->info;
GstMemory *mem;
GstFlowReturn ret;
ret = gst_hip_pool_allocator_acquire_memory (priv->alloc, &mem);
if (ret != GST_FLOW_OK) {
GST_ERROR_OBJECT (self, "Couldn't acquire memory");
return ret;
}
auto buf = gst_buffer_new ();
gst_buffer_append_memory (buf, mem);
auto hmem = GST_HIP_MEMORY_CAST (mem);
gst_buffer_add_video_meta_full (buf, GST_VIDEO_FRAME_FLAG_NONE,
GST_VIDEO_INFO_FORMAT (info), GST_VIDEO_INFO_WIDTH (info),
GST_VIDEO_INFO_HEIGHT (info), GST_VIDEO_INFO_N_PLANES (info),
hmem->info.offset, hmem->info.stride);
*buffer = buf;
return GST_FLOW_OK;
}
static gboolean
gst_hip_buffer_pool_start (GstBufferPool * pool)
{
auto self = GST_HIP_BUFFER_POOL (pool);
auto priv = self->priv;
if (!gst_hip_allocator_set_active (GST_HIP_ALLOCATOR (priv->alloc), TRUE)) {
GST_ERROR_OBJECT (self, "Couldn't activate allocator");
return FALSE;
}
return TRUE;
}
static gboolean
gst_hip_buffer_pool_stop (GstBufferPool * pool)
{
auto self = GST_HIP_BUFFER_POOL (pool);
auto priv = self->priv;
if (priv->alloc)
gst_hip_allocator_set_active (GST_HIP_ALLOCATOR (priv->alloc), FALSE);
return GST_BUFFER_POOL_CLASS (parent_class)->stop (pool);
}
GstBufferPool *
gst_hip_buffer_pool_new (GstHipDevice * device)
{
g_return_val_if_fail (GST_IS_HIP_DEVICE (device), nullptr);
auto self = (GstHipBufferPool *)
g_object_new (GST_TYPE_HIP_BUFFER_POOL, nullptr);
gst_object_ref_sink (self);
self->device = (GstHipDevice *) gst_object_ref (device);
return GST_BUFFER_POOL_CAST (self);
}

55
subprojects/gst-plugins-bad/sys/hip/gsthipbufferpool.h Normal file
View File

@ -0,0 +1,55 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include <gst/video/video.h>
#include "gsthip_fwd.h"
G_BEGIN_DECLS
#define GST_TYPE_HIP_BUFFER_POOL (gst_hip_buffer_pool_get_type ())
#define GST_HIP_BUFFER_POOL(obj) (G_TYPE_CHECK_INSTANCE_CAST ((obj),GST_TYPE_HIP_BUFFER_POOL,GstHipBufferPool))
#define GST_HIP_BUFFER_POOL_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST ((klass), GST_TYPE_HIP_BUFFER_POOL,GstHipBufferPoolClass))
#define GST_HIP_BUFFER_POOL_GET_CLASS(obj) (G_TYPE_INSTANCE_GET_CLASS((obj), GST_TYPE_HIP_BUFFER_POOL,GstHipBufferPoolClass))
#define GST_IS_HIP_BUFFER_POOL(obj) (G_TYPE_CHECK_INSTANCE_TYPE ((obj),GST_TYPE_HIP_BUFFER_POOL))
#define GST_IS_HIP_BUFFER_POOL_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE ((klass), GST_TYPE_HIP_BUFFER_POOL))
#define GST_HIP_BUFFER_POOL_CAST(obj) ((GstHipBufferPool*)(obj))
struct _GstHipBufferPool
{
GstBufferPool parent;
GstHipDevice *device;
GstHipBufferPoolPrivate *priv;
};
struct _GstHipBufferPoolClass
{
GstBufferPoolClass parent_class;
};
GType gst_hip_buffer_pool_get_type (void);
GstBufferPool * gst_hip_buffer_pool_new (GstHipDevice * device);
G_END_DECLS

1785
subprojects/gst-plugins-bad/sys/hip/gsthipconverter.cpp Normal file
View File

File diff suppressed because it is too large Load Diff

68
subprojects/gst-plugins-bad/sys/hip/gsthipconverter.h Normal file
View File

@ -0,0 +1,68 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include "gsthip.h"
G_BEGIN_DECLS
#define GST_TYPE_HIP_CONVERTER (gst_hip_converter_get_type())
#define GST_HIP_CONVERTER(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj),GST_TYPE_HIP_CONVERTER,GstHipConverter))
#define GST_HIP_CONVERTER_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass),GST_TYPE_HIP_CONVERTER,GstHipConverterClass))
#define GST_HIP_CONVERTER_GET_CLASS(obj) (GST_HIP_CONVERTER_CLASS(G_OBJECT_GET_CLASS(obj)))
#define GST_IS_HIP_CONVERTER(obj) (G_TYPE_CHECK_INSTANCE_TYPE((obj),GST_TYPE_HIP_CONVERTER))
#define GST_IS_HIP_CONVERTER_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE((klass),GST_TYPE_HIP_CONVERTER))
#define GST_HIP_CONVERTER_CAST(obj) ((GstHipConverter*)(obj))
typedef struct _GstHipConverter GstHipConverter;
typedef struct _GstHipConverterClass GstHipConverterClass;
typedef struct _GstHipConverterPrivate GstHipConverterPrivate;
struct _GstHipConverter
{
GstObject parent;
GstHipDevice *device;
/*< private >*/
GstHipConverterPrivate *priv;
gpointer _gst_reserved[GST_PADDING];
};
struct _GstHipConverterClass
{
GstObjectClass parent_class;
/*< private >*/
gpointer _gst_reserved[GST_PADDING];
};
GType gst_hip_converter_get_type (void);
GstHipConverter * gst_hip_converter_new (GstHipDevice * device,
const GstVideoInfo * in_info,
const GstVideoInfo * out_info,
GstStructure * config);
gboolean gst_hip_converter_convert_frame (GstHipConverter * converter,
GstBuffer * in_buf,
GstBuffer * out_buf);
G_END_DECLS

1868
subprojects/gst-plugins-bad/sys/hip/gsthipconvertscale.cpp Normal file
View File

File diff suppressed because it is too large Load Diff

66
subprojects/gst-plugins-bad/sys/hip/gsthipconvertscale.h Normal file
View File

@ -0,0 +1,66 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include "gsthipbasefilter.h"
G_BEGIN_DECLS
#define GST_TYPE_HIP_BASE_CONVERT (gst_hip_base_convert_get_type())
#define GST_HIP_BASE_CONVERT(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj),GST_TYPE_HIP_BASE_CONVERT,GstHipBaseConvert))
#define GST_HIP_BASE_CONVERT_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass), GST_TYPE_HIP_BASE_CONVERT,GstHipBaseConvertClass))
#define GST_HIP_BASE_CONVERT_GET_CLASS(obj) (G_TYPE_INSTANCE_GET_CLASS((obj), GST_TYPE_HIP_BASE_CONVERT,GstHipBaseConvertClass))
#define GST_IS_HIP_BASE_CONVERT(obj) (G_TYPE_CHECK_INSTANCE_TYPE((obj),GST_TYPE_HIP_BASE_CONVERT))
#define GST_IS_HIP_BASE_CONVERT_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE((klass), GST_TYPE_HIP_BASE_CONVERT))
typedef struct _GstHipBaseConvert GstHipBaseConvert;
typedef struct _GstHipBaseConvertClass GstHipBaseConvertClass;
typedef struct _GstHipBaseConvertPrivate GstHipBaseConvertPrivate;
struct _GstHipBaseConvert
{
GstHipBaseFilter parent;
GstHipBaseConvertPrivate *priv;
};
struct _GstHipBaseConvertClass
{
GstHipBaseFilter parent_class;
};
GType gst_hip_base_convert_get_type (void);
G_DEFINE_AUTOPTR_CLEANUP_FUNC(GstHipBaseConvert, gst_object_unref)
#define GST_TYPE_HIP_CONVERT_SCALE (gst_hip_convert_scale_get_type())
G_DECLARE_FINAL_TYPE (GstHipConvertScale, gst_hip_convert_scale,
GST, HIP_CONVERT_SCALE, GstHipBaseConvert)
#define GST_TYPE_HIP_CONVERT (gst_hip_convert_get_type())
G_DECLARE_FINAL_TYPE (GstHipConvert, gst_hip_convert,
GST, HIP_CONVERT, GstHipBaseConvert)
#define GST_TYPE_HIP_SCALE (gst_hip_scale_get_type())
G_DECLARE_FINAL_TYPE (GstHipScale, gst_hip_scale,
GST, HIP_SCALE, GstHipBaseConvert)
G_END_DECLS

216
subprojects/gst-plugins-bad/sys/hip/gsthipdevice.cpp Normal file
View File

@ -0,0 +1,216 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include "config.h"
#endif
#include "gsthip.h"
#include <mutex>
#ifndef GST_DISABLE_GST_DEBUG
#define GST_CAT_DEFAULT ensure_debug_category()
static GstDebugCategory *
ensure_debug_category (void)
{
static GstDebugCategory *cat = nullptr;
static std::once_flag once;
std::call_once (once,[&] {
cat = _gst_debug_category_new ("hipdevice", 0, "hipdevice");
});
return cat;
}
#endif
enum
{
PROP_0,
PROP_DEVICE_ID,
PROP_TEXTURE2D_SUPPORT,
};
struct _GstHipDevicePrivate
{
guint device_id;
gboolean texture_support;
};
#define gst_hip_device_parent_class parent_class
G_DEFINE_TYPE (GstHipDevice, gst_hip_device, GST_TYPE_OBJECT);
static void gst_hip_device_get_property (GObject * object, guint prop_id,
GValue * value, GParamSpec * pspec);
static void gst_hip_device_finalize (GObject * object);
static void
gst_hip_device_class_init (GstHipDeviceClass * klass)
{
auto object_class = G_OBJECT_CLASS (klass);
object_class->get_property = gst_hip_device_get_property;
object_class->finalize = gst_hip_device_finalize;
g_object_class_install_property (object_class, PROP_DEVICE_ID,
g_param_spec_uint ("device-id", "Device ID", "Device ID",
0, G_MAXUINT, 0,
(GParamFlags) (G_PARAM_READABLE | G_PARAM_STATIC_STRINGS)));
g_object_class_install_property (object_class, PROP_TEXTURE2D_SUPPORT,
g_param_spec_boolean ("texture2d-support", "Texture2D support",
"Texture2D support", FALSE,
(GParamFlags) (G_PARAM_READABLE | G_PARAM_STATIC_STRINGS)));
gst_hip_memory_init_once ();
}
static void
gst_hip_device_init (GstHipDevice * self)
{
self->priv = new GstHipDevicePrivate ();
}
static void
gst_hip_device_get_property (GObject * object, guint prop_id,
GValue * value, GParamSpec * pspec)
{
auto self = GST_HIP_DEVICE (object);
auto priv = self->priv;
switch (prop_id) {
case PROP_DEVICE_ID:
g_value_set_uint (value, priv->device_id);
break;
case PROP_TEXTURE2D_SUPPORT:
g_value_set_boolean (value, priv->texture_support);
break;
default:
G_OBJECT_WARN_INVALID_PROPERTY_ID (object, prop_id, pspec);
break;
}
}
static void
gst_hip_device_finalize (GObject * object)
{
auto self = GST_HIP_DEVICE (object);
delete self->priv;
G_OBJECT_CLASS (parent_class)->finalize (object);
}
static hipError_t
gst_hip_init_once (void)
{
static hipError_t ret = hipErrorInitializationError;
static std::once_flag once;
std::call_once (once,[&] {
ret = hipInit (0);
});
return ret;
}
GstHipDevice *
gst_hip_device_new (guint device_id)
{
auto hip_ret = gst_hip_init_once ();
if (hip_ret != hipSuccess) {
GST_DEBUG ("Couldn't initialize HIP, error: %d", hip_ret);
return nullptr;
}
int num_dev = 0;
hip_ret = hipGetDeviceCount (&num_dev);
if (hip_ret != hipSuccess || num_dev <= 0) {
GST_DEBUG ("No supported HIP device, error: %d", hip_ret);
return nullptr;
}
if ((guint) num_dev <= device_id) {
GST_DEBUG ("Num device %d <= requested device id %d", num_dev, device_id);
return nullptr;
}
gboolean texture_support = FALSE;
int val = 0;
hip_ret = hipDeviceGetAttribute (&val,
hipDeviceAttributeMaxTexture2DWidth, device_id);
if (hip_ret == hipSuccess && val > 0) {
hip_ret = hipDeviceGetAttribute (&val,
hipDeviceAttributeMaxTexture2DHeight, device_id);
if (hip_ret == hipSuccess && val > 0) {
hip_ret = hipDeviceGetAttribute (&val,
hipDeviceAttributeTextureAlignment, device_id);
if (hip_ret == hipSuccess && val > 0) {
texture_support = TRUE;
}
}
}
auto self = (GstHipDevice *) g_object_new (GST_TYPE_HIP_DEVICE, nullptr);
gst_object_ref_sink (self);
self->priv->device_id = device_id;
self->priv->texture_support = texture_support;
return self;
}
gboolean
gst_hip_device_set_current (GstHipDevice * device)
{
g_return_val_if_fail (GST_IS_HIP_DEVICE (device), FALSE);
auto hip_ret = hipSetDevice (device->priv->device_id);
if (!gst_hip_result (hip_ret)) {
GST_ERROR_OBJECT (device, "hipSetDevice result %d", hip_ret);
return FALSE;
}
return TRUE;
}
hipError_t
gst_hip_device_get_attribute (GstHipDevice * device, hipDeviceAttribute_t attr,
gint * value)
{
g_return_val_if_fail (GST_IS_HIP_DEVICE (device), hipErrorInvalidDevice);
return hipDeviceGetAttribute (value, attr, device->priv->device_id);
}
gboolean
gst_hip_device_is_equal (GstHipDevice * device1, GstHipDevice * device2)
{
if (!device1 || !device2)
return FALSE;
g_return_val_if_fail (GST_IS_HIP_DEVICE (device1), FALSE);
g_return_val_if_fail (GST_IS_HIP_DEVICE (device2), FALSE);
if (device1 == device2)
return TRUE;
if (device1->priv->device_id == device2->priv->device_id)
return TRUE;
return FALSE;
}

64
subprojects/gst-plugins-bad/sys/hip/gsthipdevice.h Normal file
View File

@ -0,0 +1,64 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include "gsthip_fwd.h"
#include <hip/hip_runtime.h>
G_BEGIN_DECLS
#define GST_TYPE_HIP_DEVICE (gst_hip_device_get_type())
#define GST_HIP_DEVICE(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj), GST_TYPE_HIP_DEVICE,GstHipDevice))
#define GST_HIP_DEVICE_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass), GST_TYPE_HIP_DEVICE,GstHipDeviceClass))
#define GST_HIP_DEVICE_GET_CLASS(obj) (G_TYPE_INSTANCE_GET_CLASS((obj), GST_TYPE_HIP_DEVICE,GstHipDeviceClass))
#define GST_IS_HIP_DEVICE(obj) (G_TYPE_CHECK_INSTANCE_TYPE ((obj),GST_TYPE_HIP_DEVICE))
#define GST_IS_HIP_DEVICE_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE ((klass), GST_TYPE_HIP_DEVICE))
#define GST_HIP_DEVICE_CONTEXT_TYPE "gst.hip.device"
struct _GstHipDevice
{
GstObject object;
/*< private >*/
GstHipDevicePrivate *priv;
};
struct _GstHipDeviceClass
{
GstObjectClass parent_class;
};
GType gst_hip_device_get_type (void);
GstHipDevice * gst_hip_device_new (guint device_id);
gboolean gst_hip_device_set_current (GstHipDevice * device);
hipError_t gst_hip_device_get_attribute (GstHipDevice * device,
hipDeviceAttribute_t attr,
gint * value);
gboolean gst_hip_device_is_equal (GstHipDevice * device1,
GstHipDevice * device2);
G_END_DECLS

1051
subprojects/gst-plugins-bad/sys/hip/gsthipmemory.cpp Normal file
View File

File diff suppressed because it is too large Load Diff

141
subprojects/gst-plugins-bad/sys/hip/gsthipmemory.h Normal file
View File

@ -0,0 +1,141 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include <gst/video/video.h>
#include "gsthip_fwd.h"
G_BEGIN_DECLS
#define GST_HIP_MEMORY_CAST(obj) ((GstHipMemory *)obj)
#define GST_TYPE_HIP_ALLOCATOR (gst_hip_allocator_get_type())
#define GST_HIP_ALLOCATOR(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj), GST_TYPE_HIP_ALLOCATOR, GstHipAllocator))
#define GST_HIP_ALLOCATOR_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass), GST_TYPE_HIP_ALLOCATOR, GstHipAllocatorClass))
#define GST_IS_HIP_ALLOCATOR(obj) (G_TYPE_CHECK_INSTANCE_TYPE((obj), GST_TYPE_HIP_ALLOCATOR))
#define GST_IS_HIP_ALLOCATOR_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE((klass), GST_TYPE_HIP_ALLOCATOR))
#define GST_HIP_ALLOCATOR_GET_CLASS(obj) (G_TYPE_INSTANCE_GET_CLASS((obj), GST_TYPE_HIP_ALLOCATOR, GstHipAllocatorClass))
#define GST_HIP_ALLOCATOR_CAST(obj) ((GstHipAllocator *)obj)
#define GST_TYPE_HIP_POOL_ALLOCATOR (gst_hip_pool_allocator_get_type())
#define GST_HIP_POOL_ALLOCATOR(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj), GST_TYPE_HIP_POOL_ALLOCATOR, GstHipPoolAllocator))
#define GST_HIP_POOL_ALLOCATOR_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass), GST_TYPE_HIP_POOL_ALLOCATOR, GstHipPoolAllocatorClass))
#define GST_IS_HIP_POOL_ALLOCATOR(obj) (G_TYPE_CHECK_INSTANCE_TYPE((obj), GST_TYPE_HIP_POOL_ALLOCATOR))
#define GST_IS_HIP_POOL_ALLOCATOR_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE((klass), GST_TYPE_HIP_POOL_ALLOCATOR))
#define GST_HIP_POOL_ALLOCATOR_GET_CLASS(obj) (G_TYPE_INSTANCE_GET_CLASS((obj), GST_TYPE_HIP_POOL_ALLOCATOR, GstHipPoolAllocatorClass))
#define GST_HIP_POOL_ALLOCATOR_CAST(obj) ((GstHipPoolAllocator *)obj)
#define GST_HIP_MEMORY_NAME "HIPMemory"
#define GST_CAPS_FEATURE_MEMORY_HIP_MEMORY "memory:HIPMemory"
#define GST_MAP_HIP ((GstMapFlags) (GST_MAP_FLAG_LAST << 1))
#define GST_MAP_READ_HIP ((GstMapFlags) (GST_MAP_READ | GST_MAP_HIP))
#define GST_MAP_WRITE_HIP ((GstMapFlags) (GST_MAP_WRITE | GST_MAP_HIP))
typedef enum
{
GST_HIP_MEMORY_TRANSFER_NEED_DOWNLOAD = (GST_MEMORY_FLAG_LAST << 0),
GST_HIP_MEMORY_TRANSFER_NEED_UPLOAD = (GST_MEMORY_FLAG_LAST << 1)
} GstHipMemoryTransfer;
struct _GstHipMemory
{
GstMemory mem;
/*< public >*/
GstHipDevice *device;
GstVideoInfo info;
/*< private >*/
GstHipMemoryPrivate *priv;
gpointer _gst_reserved[GST_PADDING];
};
void gst_hip_memory_init_once (void);
gboolean gst_is_hip_memory (GstMemory * mem);
gboolean gst_hip_memory_get_texture (GstHipMemory * mem,
guint plane,
guint filter_mode,
guint address_mode,
hipTextureObject_t * texture);
struct _GstHipAllocator
{
GstAllocator allocator;
/*< private >*/
GstHipAllocatorPrivate *priv;
gpointer _gst_reserved[GST_PADDING];
};
struct _GstHipAllocatorClass
{
GstAllocatorClass allocator_class;
gboolean (*set_active) (GstHipAllocator * allocator,
gboolean active);
/*< private >*/
gpointer _gst_reserved[GST_PADDING_LARGE];
};
GType gst_hip_allocator_get_type (void);
GstMemory * gst_hip_allocator_alloc (GstHipAllocator * allocator,
GstHipDevice * device,
const GstVideoInfo * info);
gboolean gst_hip_allocator_set_active (GstHipAllocator * allocator,
gboolean active);
struct _GstHipPoolAllocator
{
GstHipAllocator parent;
GstHipDevice *device;
GstVideoInfo info;
/*< private >*/
GstHipPoolAllocatorPrivate *priv;
gpointer _gst_reserved[GST_PADDING];
};
struct _GstHipPoolAllocatorClass
{
GstHipAllocatorClass parent_class;
/*< private >*/
gpointer _gst_reserved[GST_PADDING];
};
GType gst_hip_pool_allocator_get_type (void);
GstHipPoolAllocator * gst_hip_pool_allocator_new (GstHipDevice * device,
const GstVideoInfo * info,
GstStructure * config);
GstFlowReturn gst_hip_pool_allocator_acquire_memory (GstHipPoolAllocator * allocator,
GstMemory ** memory);
G_END_DECLS

427
subprojects/gst-plugins-bad/sys/hip/gsthipmemorycopy.cpp Normal file
View File

@ -0,0 +1,427 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include <config.h>
#endif
#include "gsthip.h"
#include "gsthipmemorycopy.h"
#include <mutex>
GST_DEBUG_CATEGORY_STATIC (gst_hip_memory_copy_debug);
#define GST_CAT_DEFAULT gst_hip_memory_copy_debug
#define GST_HIP_FORMATS \
"{ I420, YV12, NV12, NV21, P010_10LE, P012_LE, P016_LE, I420_10LE, I420_12LE, Y444, " \
"Y444_10LE, Y444_12LE, Y444_16LE, BGRA, RGBA, RGBx, BGRx, ARGB, ABGR, RGB, " \
"BGR, BGR10A2_LE, RGB10A2_LE, Y42B, I422_10LE, I422_12LE, YUY2, UYVY, RGBP, " \
"BGRP, GBR, GBR_10LE, GBR_12LE, GBR_16LE, GBRA, VUYA }"
static GstStaticPadTemplate sink_template =
GST_STATIC_PAD_TEMPLATE ("sink", GST_PAD_SINK, GST_PAD_ALWAYS,
GST_STATIC_CAPS (GST_VIDEO_CAPS_MAKE_WITH_FEATURES
(GST_CAPS_FEATURE_MEMORY_HIP_MEMORY, GST_HIP_FORMATS) "; "
GST_VIDEO_CAPS_MAKE_WITH_FEATURES
(GST_CAPS_FEATURE_MEMORY_HIP_MEMORY ","
GST_CAPS_FEATURE_META_GST_VIDEO_OVERLAY_COMPOSITION,
GST_HIP_FORMATS) "; "
GST_VIDEO_CAPS_MAKE (GST_HIP_FORMATS) "; "
GST_VIDEO_CAPS_MAKE_WITH_FEATURES
(GST_CAPS_FEATURE_MEMORY_SYSTEM_MEMORY ","
GST_CAPS_FEATURE_META_GST_VIDEO_OVERLAY_COMPOSITION,
GST_HIP_FORMATS)));
static GstStaticPadTemplate src_template =
GST_STATIC_PAD_TEMPLATE ("src", GST_PAD_SRC, GST_PAD_ALWAYS,
GST_STATIC_CAPS (GST_VIDEO_CAPS_MAKE_WITH_FEATURES
(GST_CAPS_FEATURE_MEMORY_HIP_MEMORY, GST_HIP_FORMATS) "; "
GST_VIDEO_CAPS_MAKE_WITH_FEATURES
(GST_CAPS_FEATURE_MEMORY_HIP_MEMORY ","
GST_CAPS_FEATURE_META_GST_VIDEO_OVERLAY_COMPOSITION,
GST_HIP_FORMATS) "; "
GST_VIDEO_CAPS_MAKE (GST_HIP_FORMATS) "; "
GST_VIDEO_CAPS_MAKE_WITH_FEATURES
(GST_CAPS_FEATURE_MEMORY_SYSTEM_MEMORY ","
GST_CAPS_FEATURE_META_GST_VIDEO_OVERLAY_COMPOSITION,
GST_HIP_FORMATS)));
struct _GstHipMemoryCopyPrivate
{
gboolean is_uploader;
std::recursive_mutex lock;
};
static void gst_hip_memory_copy_finalize (GObject * object);
static GstCaps *gst_hip_memory_copy_transform_caps (GstBaseTransform * trans,
GstPadDirection direction, GstCaps * caps, GstCaps * filter);
static gboolean gst_hip_memory_copy_propose_allocation (GstBaseTransform *
trans, GstQuery * decide_query, GstQuery * query);
static gboolean gst_hip_memory_copy_decide_allocation (GstBaseTransform *
trans, GstQuery * query);
static GstFlowReturn gst_hip_memory_copy_transform (GstBaseTransform * trans,
GstBuffer * inbuf, GstBuffer * outbuf);
#define gst_hip_memory_copy_parent_class parent_class
G_DEFINE_ABSTRACT_TYPE (GstHipMemoryCopy, gst_hip_memory_copy,
GST_TYPE_HIP_BASE_FILTER);
static void
gst_hip_memory_copy_class_init (GstHipMemoryCopyClass * klass)
{
auto object_class = G_OBJECT_CLASS (klass);
auto element_class = GST_ELEMENT_CLASS (klass);
auto trans_class = GST_BASE_TRANSFORM_CLASS (klass);
object_class->finalize = gst_hip_memory_copy_finalize;
gst_element_class_add_static_pad_template (element_class, &sink_template);
gst_element_class_add_static_pad_template (element_class, &src_template);
trans_class->passthrough_on_same_caps = TRUE;
trans_class->transform_caps =
GST_DEBUG_FUNCPTR (gst_hip_memory_copy_transform_caps);
trans_class->propose_allocation =
GST_DEBUG_FUNCPTR (gst_hip_memory_copy_propose_allocation);
trans_class->decide_allocation =
GST_DEBUG_FUNCPTR (gst_hip_memory_copy_decide_allocation);
trans_class->transform = GST_DEBUG_FUNCPTR (gst_hip_memory_copy_transform);
gst_type_mark_as_plugin_api (GST_TYPE_HIP_MEMORY_COPY, (GstPluginAPIFlags) 0);
GST_DEBUG_CATEGORY_INIT (gst_hip_memory_copy_debug,
"hipmemorycopy", 0, "hipmemorycopy");
}
static void
gst_hip_memory_copy_init (GstHipMemoryCopy * self)
{
self->priv = new GstHipMemoryCopyPrivate ();
}
static void
gst_hip_memory_copy_finalize (GObject * object)
{
auto self = GST_HIP_MEMORY_COPY (object);
delete self->priv;
G_OBJECT_CLASS (parent_class)->finalize (object);
}
static GstCaps *
_set_caps_features (const GstCaps * caps, const gchar * feature_name)
{
GstCaps *tmp = gst_caps_copy (caps);
guint n = gst_caps_get_size (tmp);
guint i = 0;
for (i = 0; i < n; i++) {
gst_caps_set_features (tmp, i,
gst_caps_features_new_single_static_str (feature_name));
}
return tmp;
}
static GstCaps *
gst_hip_memory_copy_transform_caps (GstBaseTransform * trans,
GstPadDirection direction, GstCaps * caps, GstCaps * filter)
{
auto self = GST_HIP_MEMORY_COPY (trans);
auto priv = self->priv;
GstCaps *result, *tmp;
GST_DEBUG_OBJECT (self,
"Transforming caps %" GST_PTR_FORMAT " in direction %s", caps,
(direction == GST_PAD_SINK) ? "sink" : "src");
if (direction == GST_PAD_SINK) {
if (priv->is_uploader) {
auto caps_hip =
_set_caps_features (caps, GST_CAPS_FEATURE_MEMORY_HIP_MEMORY);
tmp = gst_caps_merge (caps_hip, gst_caps_ref (caps));
} else {
auto caps_sys =
_set_caps_features (caps, GST_CAPS_FEATURE_MEMORY_SYSTEM_MEMORY);
tmp = gst_caps_merge (caps_sys, gst_caps_ref (caps));
}
} else {
if (priv->is_uploader) {
auto caps_sys =
_set_caps_features (caps, GST_CAPS_FEATURE_MEMORY_SYSTEM_MEMORY);
tmp = gst_caps_merge (caps_sys, gst_caps_ref (caps));
} else {
auto caps_hip =
_set_caps_features (caps, GST_CAPS_FEATURE_MEMORY_HIP_MEMORY);
tmp = gst_caps_merge (caps_hip, gst_caps_ref (caps));
}
}
if (filter) {
result = gst_caps_intersect_full (filter, tmp, GST_CAPS_INTERSECT_FIRST);
gst_caps_unref (tmp);
} else {
result = tmp;
}
GST_DEBUG_OBJECT (trans, "returning caps: %" GST_PTR_FORMAT, result);
return result;
}
static gboolean
gst_hip_memory_copy_propose_allocation (GstBaseTransform * trans,
GstQuery * decide_query, GstQuery * query)
{
auto filter = GST_HIP_BASE_FILTER (trans);
auto self = GST_HIP_MEMORY_COPY (trans);
GstVideoInfo info;
GstBufferPool *pool = nullptr;
GstCaps *caps;
guint size;
bool is_hip = false;
if (!GST_BASE_TRANSFORM_CLASS (parent_class)->propose_allocation (trans,
decide_query, query))
return FALSE;
/* passthrough, we're done */
if (!decide_query)
return TRUE;
gst_query_parse_allocation (query, &caps, nullptr);
if (!caps) {
GST_WARNING_OBJECT (self, "Allocation query without caps");
return FALSE;
}
if (!gst_video_info_from_caps (&info, caps)) {
GST_ERROR_OBJECT (self, "Invalid caps %" GST_PTR_FORMAT, caps);
return FALSE;
}
if (gst_query_get_n_allocation_pools (query) == 0) {
auto features = gst_caps_get_features (caps, 0);
if (features && gst_caps_features_contains (features,
GST_CAPS_FEATURE_MEMORY_HIP_MEMORY)) {
GST_DEBUG_OBJECT (self, "upstream support hip memory");
pool = gst_hip_buffer_pool_new (filter->device);
is_hip = true;
} else {
pool = gst_video_buffer_pool_new ();
}
auto config = gst_buffer_pool_get_config (pool);
gst_buffer_pool_config_add_option (config,
GST_BUFFER_POOL_OPTION_VIDEO_META);
if (!is_hip) {
gst_buffer_pool_config_add_option (config,
GST_BUFFER_POOL_OPTION_VIDEO_ALIGNMENT);
}
size = GST_VIDEO_INFO_SIZE (&info);
gst_buffer_pool_config_set_params (config, caps, size, 0, 0);
if (!gst_buffer_pool_set_config (pool, config)) {
GST_ERROR_OBJECT (self, "Bufferpool config failed");
gst_object_unref (pool);
return FALSE;
}
config = gst_buffer_pool_get_config (pool);
gst_buffer_pool_config_get_params (config,
nullptr, &size, nullptr, nullptr);
gst_structure_free (config);
gst_query_add_allocation_pool (query, pool, size, 0, 0);
gst_object_unref (pool);
}
gst_query_add_allocation_meta (query, GST_VIDEO_META_API_TYPE, nullptr);
gst_query_add_allocation_meta (query,
GST_VIDEO_OVERLAY_COMPOSITION_META_API_TYPE, nullptr);
return TRUE;
}
static gboolean
gst_hip_memory_copy_decide_allocation (GstBaseTransform * trans,
GstQuery * query)
{
auto filter = GST_HIP_BASE_FILTER (trans);
auto self = GST_HIP_MEMORY_COPY (trans);
GstBufferPool *pool = nullptr;
GstVideoInfo info;
guint min, max, size;
GstCaps *caps = nullptr;
bool update_pool = false;
gst_query_parse_allocation (query, &caps, nullptr);
if (!caps) {
GST_WARNING_OBJECT (self, "Allocation query without caps");
return FALSE;
}
if (!gst_video_info_from_caps (&info, caps)) {
GST_ERROR_OBJECT (self, "Invalid caps %" GST_PTR_FORMAT, caps);
return FALSE;
}
if (gst_query_get_n_allocation_pools (query) > 0) {
gst_query_parse_nth_allocation_pool (query, 0, &pool, &size, &min, &max);
update_pool = true;
} else {
size = info.size;
min = max = 0;
}
auto features = gst_caps_get_features (caps, 0);
if (features && gst_caps_features_contains (features,
GST_CAPS_FEATURE_MEMORY_HIP_MEMORY)) {
GST_DEBUG_OBJECT (self, "upstream support hip memory");
if (pool) {
if (!GST_IS_HIP_BUFFER_POOL (pool)) {
gst_clear_object (&pool);
} else {
auto hpool = GST_HIP_BUFFER_POOL (pool);
if (!gst_hip_device_is_equal (hpool->device, filter->device))
gst_clear_object (&pool);
}
}
if (!pool)
pool = gst_hip_buffer_pool_new (filter->device);
} else {
pool = gst_video_buffer_pool_new ();
}
auto config = gst_buffer_pool_get_config (pool);
gst_buffer_pool_config_add_option (config, GST_BUFFER_POOL_OPTION_VIDEO_META);
gst_buffer_pool_config_set_params (config, caps, size, min, max);
gst_buffer_pool_set_config (pool, config);
config = gst_buffer_pool_get_config (pool);
gst_buffer_pool_config_get_params (config, nullptr, &size, nullptr, nullptr);
gst_structure_free (config);
if (update_pool)
gst_query_set_nth_allocation_pool (query, 0, pool, size, min, max);
else
gst_query_add_allocation_pool (query, pool, size, min, max);
gst_object_unref (pool);
return GST_BASE_TRANSFORM_CLASS (parent_class)->decide_allocation (trans,
query);
}
static GstFlowReturn
gst_hip_memory_copy_system_copy (GstHipMemoryCopy * self,
GstBuffer * inbuf, GstBuffer * outbuf)
{
auto filter = GST_HIP_BASE_FILTER (self);
GstVideoFrame in_frame, out_frame;
GstFlowReturn ret = GST_FLOW_OK;
if (!gst_video_frame_map (&in_frame, &filter->in_info, inbuf, GST_MAP_READ)) {
GST_ERROR_OBJECT (self, "Couldn't map input frame");
return GST_FLOW_ERROR;
}
if (!gst_video_frame_map (&out_frame, &filter->out_info, outbuf,
GST_MAP_WRITE)) {
GST_ERROR_OBJECT (self, "Couldn't map output frame");
gst_video_frame_unmap (&in_frame);
return GST_FLOW_ERROR;
}
if (!gst_video_frame_copy (&out_frame, &in_frame)) {
GST_ERROR_OBJECT (self, "Copy failed");
ret = GST_FLOW_ERROR;
}
gst_video_frame_unmap (&out_frame);
gst_video_frame_unmap (&in_frame);
return ret;
}
static GstFlowReturn
gst_hip_memory_copy_transform (GstBaseTransform * trans, GstBuffer * inbuf,
GstBuffer * outbuf)
{
auto self = GST_HIP_MEMORY_COPY (trans);
return gst_hip_memory_copy_system_copy (self, inbuf, outbuf);
}
struct _GstHipUpload
{
GstHipMemoryCopy parent;
};
G_DEFINE_TYPE (GstHipUpload, gst_hip_upload, GST_TYPE_HIP_MEMORY_COPY);
static void
gst_hip_upload_class_init (GstHipUploadClass * klass)
{
auto element_class = GST_ELEMENT_CLASS (klass);
gst_element_class_set_static_metadata (element_class,
"HIP Uploader", "Filter/Video",
"Uploads system memory into HIP device memory",
"Seungha Yang <seungha@centricular.com>");
}
static void
gst_hip_upload_init (GstHipUpload * self)
{
auto memcpy = GST_HIP_MEMORY_COPY (self);
memcpy->priv->is_uploader = true;
}
struct _GstHipDownload
{
GstHipMemoryCopy parent;
};
G_DEFINE_TYPE (GstHipDownload, gst_hip_download, GST_TYPE_HIP_MEMORY_COPY);
static void
gst_hip_download_class_init (GstHipDownloadClass * klass)
{
auto element_class = GST_ELEMENT_CLASS (klass);
gst_element_class_set_static_metadata (element_class,
"HIP Downloader", "Filter/Video",
"Downloads HIP device memory into system memory",
"Seungha Yang <seungha@centricular.com>");
}
static void
gst_hip_download_init (GstHipDownload * self)
{
auto memcpy = GST_HIP_MEMORY_COPY (self);
memcpy->priv->is_uploader = false;
}

62
subprojects/gst-plugins-bad/sys/hip/gsthipmemorycopy.h Normal file
View File

@ -0,0 +1,62 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include "gsthipbasefilter.h"
G_BEGIN_DECLS
#define GST_TYPE_HIP_MEMORY_COPY (gst_hip_memory_copy_get_type())
#define GST_HIP_MEMORY_COPY(obj) (G_TYPE_CHECK_INSTANCE_CAST((obj),GST_TYPE_HIP_MEMORY_COPY,GstHipMemoryCopy))
#define GST_HIP_MEMORY_COPY_CLASS(klass) (G_TYPE_CHECK_CLASS_CAST((klass),GST_TYPE_HIP_MEMORY_COPY,GstHipMemoryCopyClass))
#define GST_HIP_MEMORY_COPY_GET_CLASS(obj) (GST_HIP_MEMORY_COPY_CLASS(G_OBJECT_GET_CLASS(obj)))
#define GST_IS_HIP_MEMORY_COPY(obj) (G_TYPE_CHECK_INSTANCE_TYPE((obj),GST_TYPE_HIP_MEMORY_COPY))
#define GST_IS_HIP_MEMORY_COPY_CLASS(klass) (G_TYPE_CHECK_CLASS_TYPE((klass),GST_TYPE_HIP_MEMORY_COPY))
typedef struct _GstHipMemoryCopy GstHipMemoryCopy;
typedef struct _GstHipMemoryCopyClass GstHipMemoryCopyClass;
typedef struct _GstHipMemoryCopyPrivate GstHipMemoryCopyPrivate;
struct _GstHipMemoryCopy
{
GstHipBaseFilter parent;
GstHipMemoryCopyPrivate *priv;
};
struct _GstHipMemoryCopyClass
{
GstHipBaseFilterClass parent_class;
};
GType gst_hip_memory_copy_get_type (void);
G_DEFINE_AUTOPTR_CLEANUP_FUNC(GstHipMemoryCopy, gst_object_unref)
#define GST_TYPE_HIP_UPLOAD (gst_hip_upload_get_type())
G_DECLARE_FINAL_TYPE (GstHipUpload,
gst_hip_upload, GST, HIP_UPLOAD, GstHipMemoryCopy);
#define GST_TYPE_HIP_DOWNLOAD (gst_hip_download_get_type())
G_DECLARE_FINAL_TYPE (GstHipDownload,
gst_hip_download, GST, HIP_DOWNLOAD, GstHipMemoryCopy);
G_END_DECLS

106
subprojects/gst-plugins-bad/sys/hip/gsthiprtc.cpp Normal file
View File

@ -0,0 +1,106 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include "config.h"
#endif
#include "gsthiprtc.h"
#include "gsthip.h"
#include <hip/hiprtc.h>
#include <mutex>
#include <vector>
#include <string>
#ifndef GST_DISABLE_GST_DEBUG
#define GST_CAT_DEFAULT ensure_debug_category()
static GstDebugCategory *
ensure_debug_category (void)
{
static GstDebugCategory *cat = nullptr;
static std::once_flag once;
std::call_once (once,[&] {
cat = _gst_debug_category_new ("hiprtc", 0, "hiprtc");
});
return cat;
}
#endif
gchar *
gst_hip_rtc_compile (GstHipDevice * device,
const gchar * source, const gchar ** options, guint num_options)
{
hiprtcProgram prog;
auto rtc_ret = hiprtcCreateProgram (&prog, source, "program.cpp",
0, nullptr, nullptr);
if (rtc_ret != HIPRTC_SUCCESS) {
GST_ERROR_OBJECT (device, "Couldn't create program, ret: %d", rtc_ret);
return nullptr;
}
guint device_id;
g_object_get (device, "device-id", &device_id, nullptr);
hipDeviceProp_t props = { };
auto hip_ret = hipGetDeviceProperties (&props, device_id);
if (!gst_hip_result (hip_ret)) {
GST_ERROR_OBJECT (device, "Couldn't query device property");
return nullptr;
}
rtc_ret = hiprtcCompileProgram (prog, num_options, options);
if (rtc_ret != HIPRTC_SUCCESS) {
size_t log_size = 0;
gchar *err_str = nullptr;
rtc_ret = hiprtcGetProgramLogSize (prog, &log_size);
if (rtc_ret == HIPRTC_SUCCESS) {
err_str = (gchar *) g_malloc0 (log_size);
err_str[log_size - 1] = '\0';
hiprtcGetProgramLog (prog, err_str);
}
GST_ERROR_OBJECT (device, "Couldn't compile program, ret: %d (%s)",
rtc_ret, GST_STR_NULL (err_str));
g_free (err_str);
return nullptr;
}
size_t code_size;
rtc_ret = hiprtcGetCodeSize (prog, &code_size);
if (rtc_ret != HIPRTC_SUCCESS) {
GST_ERROR_OBJECT (device, "Couldn't get code size, ret: %d", rtc_ret);
return nullptr;
}
auto code = (gchar *) g_malloc0 (code_size);
rtc_ret = hiprtcGetCode (prog, code);
if (rtc_ret != HIPRTC_SUCCESS) {
GST_ERROR_OBJECT (device, "Couldn't get code, ret: %d", rtc_ret);
g_free (code);
return nullptr;
}
hiprtcDestroyProgram (&prog);
return code;
}

32
subprojects/gst-plugins-bad/sys/hip/gsthiprtc.h Normal file
View File

@ -0,0 +1,32 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include "gsthipdevice.h"
G_BEGIN_DECLS
gchar * gst_hip_rtc_compile (GstHipDevice * device,
const gchar * source,
const gchar ** options,
guint num_options);
G_END_DECLS

249
subprojects/gst-plugins-bad/sys/hip/gsthiputils.cpp Normal file
View File

@ -0,0 +1,249 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include "config.h"
#endif
#include "gsthip.h"
#include <mutex>
#ifndef GST_DISABLE_GST_DEBUG
#define GST_CAT_DEFAULT ensure_debug_category()
static GstDebugCategory *
ensure_debug_category (void)
{
static GstDebugCategory *cat = nullptr;
static std::once_flag once;
std::call_once (once,[&] {
cat = _gst_debug_category_new ("hiputils", 0, "hiputils");
});
return cat;
}
#endif
gboolean
_gst_hip_result (hipError_t result, GstDebugCategory * cat, const gchar * file,
const gchar * function, gint line)
{
if (result != hipSuccess) {
#ifndef GST_DISABLE_GST_DEBUG
auto error_name = hipGetErrorName (result);
auto error_str = hipGetErrorString (result);
gst_debug_log (cat, GST_LEVEL_ERROR, file, function, line,
NULL, "HIP call failed: %s, %s", error_name, error_str);
#endif
return FALSE;
}
return TRUE;
}
static void
context_set_hip_device (GstContext * context, GstHipDevice * device)
{
g_return_if_fail (context != nullptr);
guint device_id;
g_object_get (device, "device-id", &device_id, nullptr);
auto s = gst_context_writable_structure (context);
gst_structure_set (s, "device", GST_TYPE_HIP_DEVICE, device,
"device-id", G_TYPE_UINT, device_id, nullptr);
}
static gboolean
pad_query (const GValue * item, GValue * value, gpointer user_data)
{
GstPad *pad = (GstPad *) g_value_get_object (item);
GstQuery *query = (GstQuery *) user_data;
gboolean res;
res = gst_pad_peer_query (pad, query);
if (res) {
g_value_set_boolean (value, TRUE);
return FALSE;
}
return TRUE;
}
static gboolean
run_query (GstElement * element, GstQuery * query, GstPadDirection direction)
{
GstIterator *it;
GstIteratorFoldFunction func = pad_query;
GValue res = G_VALUE_INIT;
g_value_init (&res, G_TYPE_BOOLEAN);
g_value_set_boolean (&res, FALSE);
/* Ask neighbor */
if (direction == GST_PAD_SRC)
it = gst_element_iterate_src_pads (element);
else
it = gst_element_iterate_sink_pads (element);
while (gst_iterator_fold (it, func, &res, query) == GST_ITERATOR_RESYNC)
gst_iterator_resync (it);
gst_iterator_free (it);
return g_value_get_boolean (&res);
}
static void
run_hip_context_query (GstElement * element, GstHipDevice ** device)
{
GstQuery *query;
GstContext *ctx = nullptr;
query = gst_query_new_context (GST_HIP_DEVICE_CONTEXT_TYPE);
if (run_query (element, query, GST_PAD_SRC)) {
gst_query_parse_context (query, &ctx);
if (ctx)
gst_element_set_context (element, ctx);
}
if (*device == nullptr && run_query (element, query, GST_PAD_SINK)) {
gst_query_parse_context (query, &ctx);
if (ctx)
gst_element_set_context (element, ctx);
}
if (*device == nullptr) {
auto msg = gst_message_new_need_context (GST_OBJECT_CAST (element),
GST_HIP_DEVICE_CONTEXT_TYPE);
gst_element_post_message (element, msg);
}
gst_query_unref (query);
}
gboolean
gst_hip_ensure_element_data (GstElement * element, gint device_id,
GstHipDevice ** device)
{
if (*device)
return TRUE;
run_hip_context_query (element, device);
if (*device)
return TRUE;
guint target_device_id = 0;
if (device_id > 0)
target_device_id = device_id;
*device = gst_hip_device_new (target_device_id);
if (*device == nullptr) {
GST_ERROR_OBJECT (element,
"Couldn't create new device with adapter index %d", target_device_id);
return FALSE;
} else {
auto ctx = gst_context_new_hip_device (*device);
gst_element_set_context (element, ctx);
auto msg = gst_message_new_have_context (GST_OBJECT_CAST (element), ctx);
gst_element_post_message (GST_ELEMENT_CAST (element), msg);
}
return TRUE;
}
gboolean
gst_hip_handle_set_context (GstElement * element, GstContext * context,
gint device_id, GstHipDevice ** device)
{
g_return_val_if_fail (GST_IS_ELEMENT (element), FALSE);
g_return_val_if_fail (device != nullptr, FALSE);
if (!context)
return FALSE;
auto context_type = gst_context_get_context_type (context);
if (g_strcmp0 (context_type, GST_HIP_DEVICE_CONTEXT_TYPE) == 0) {
GstHipDevice *other_device = nullptr;
guint other_idx = 0;
/* If we had device already, will not replace it */
if (*device)
return TRUE;
auto s = gst_context_get_structure (context);
if (gst_structure_get (s, "device", GST_TYPE_HIP_DEVICE, &other_device,
"device-id", G_TYPE_UINT, &other_idx, nullptr)) {
if (device_id == -1 || (guint) device_id == other_idx) {
*device = other_device;
return TRUE;
}
gst_object_unref (other_device);
}
}
return FALSE;
}
gboolean
gst_hip_handle_context_query (GstElement * element, GstQuery * query,
GstHipDevice * device)
{
const gchar *context_type;
GstContext *context;
g_return_val_if_fail (GST_IS_ELEMENT (element), FALSE);
g_return_val_if_fail (GST_IS_QUERY (query), FALSE);
if (!GST_IS_HIP_DEVICE (device))
return FALSE;
gst_query_parse_context_type (query, &context_type);
if (g_strcmp0 (context_type, GST_HIP_DEVICE_CONTEXT_TYPE) != 0)
return FALSE;
GstContext *old_ctx = nullptr;
gst_query_parse_context (query, &old_ctx);
if (old_ctx)
context = gst_context_copy (old_ctx);
else
context = gst_context_new (GST_HIP_DEVICE_CONTEXT_TYPE, TRUE);
context_set_hip_device (context, device);
gst_query_set_context (query, context);
gst_context_unref (context);
GST_DEBUG_OBJECT (element, "successfully set %" GST_PTR_FORMAT
" on %" GST_PTR_FORMAT, device, query);
return TRUE;
}
GstContext *
gst_context_new_hip_device (GstHipDevice * device)
{
g_return_val_if_fail (GST_HIP_DEVICE (device), nullptr);
auto ctx = gst_context_new (GST_HIP_DEVICE_CONTEXT_TYPE, TRUE);
context_set_hip_device (ctx, device);
return ctx;
}

57
subprojects/gst-plugins-bad/sys/hip/gsthiputils.h Normal file
View File

@ -0,0 +1,57 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#pragma once
#include <gst/gst.h>
#include "gsthip_fwd.h"
G_BEGIN_DECLS
gboolean _gst_hip_result (hipError_t result,
GstDebugCategory * cat,
const gchar * file,
const gchar * function,
gint line);
#ifndef GST_DISABLE_GST_DEBUG
#define gst_hip_result(result) \
_gst_hip_result(result, GST_CAT_DEFAULT, __FILE__, GST_FUNCTION, __LINE__)
#else
#define gst_hip_result(result) \
_gst_hip_result(result, NULL, __FILE__, GST_FUNCTION, __LINE__)
#endif /* GST_DISABLE_GST_DEBUG */
gboolean gst_hip_ensure_element_data (GstElement * element,
gint device_id,
GstHipDevice ** device);
gboolean gst_hip_handle_set_context (GstElement * element,
GstContext * context,
gint device_id,
GstHipDevice ** device);
gboolean gst_hip_handle_context_query (GstElement * element,
GstQuery * query,
GstHipDevice * device);
GstContext * gst_context_new_hip_device (GstHipDevice * device);
G_END_DECLS

168
subprojects/gst-plugins-bad/sys/hip/kernel/converter-unpack.cu Normal file
View File

@ -0,0 +1,168 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#if defined(__NVCC__) || defined(__HIPCC__)
extern "C" {
__global__ void
GstHipConverterUnpack_RGB_RGBx
(unsigned char *src, unsigned char *dst, int width, int height,
int src_stride, int dst_stride)
{
int x_pos = blockIdx.x * blockDim.x + threadIdx.x;
int y_pos = blockIdx.y * blockDim.y + threadIdx.y;
if (x_pos < width && y_pos < height) {
int dst_pos = x_pos * 4 + y_pos * dst_stride;
int src_pos = x_pos * 3 + y_pos * src_stride;
dst[dst_pos] = src[src_pos];
dst[dst_pos + 1] = src[src_pos + 1];
dst[dst_pos + 2] = src[src_pos + 2];
dst[dst_pos + 3] = 0xff;
}
}
__global__ void
GstHipConverterUnpack_RGB10A2_ARGB64
(unsigned char *src, unsigned char *dst, int width, int height,
int src_stride, int dst_stride)
{
int x_pos = blockIdx.x * blockDim.x + threadIdx.x;
int y_pos = blockIdx.y * blockDim.y + threadIdx.y;
if (x_pos < width && y_pos < height) {
unsigned short a, r, g, b;
unsigned int val;
int dst_pos = x_pos * 8 + y_pos * dst_stride;
val = *(unsigned int *)&src[x_pos * 4 + y_pos * src_stride];
a = (val >> 30) & 0x03;
a = (a << 14) | (a << 12) | (a << 10) | (a << 8) | (a << 6) | (a << 4) | (a << 2) | (a << 0);
r = (val & 0x3ff);
r = (r << 6) | (r >> 4);
g = ((val >> 10) & 0x3ff);
g = (g << 6) | (g >> 4);
b = ((val >> 20) & 0x3ff);
b = (b << 6) | (b >> 4);
*(unsigned short *) &dst[dst_pos] = a;
*(unsigned short *) &dst[dst_pos + 2] = r;
*(unsigned short *) &dst[dst_pos + 4] = g;
*(unsigned short *) &dst[dst_pos + 6] = b;
}
}
__global__ void
GstHipConverterUnpack_BGR10A2_ARGB64
(unsigned char *src, unsigned char *dst, int width, int height,
int src_stride, int dst_stride)
{
int x_pos = blockIdx.x * blockDim.x + threadIdx.x;
int y_pos = blockIdx.y * blockDim.y + threadIdx.y;
if (x_pos < width && y_pos < height) {
unsigned short a, r, g, b;
unsigned int val;
int dst_pos = x_pos * 8 + y_pos * dst_stride;
val = *(unsigned int *)&src[x_pos * 4 + y_pos * src_stride];
a = (val >> 30) & 0x03;
a = (a << 14) | (a << 12) | (a << 10) | (a << 8) | (a << 6) | (a << 4) | (a << 2) | (a << 0);
b = (val & 0x3ff);
b = (b << 6) | (b >> 4);
g = ((val >> 10) & 0x3ff);
g = (g << 6) | (g >> 4);
r = ((val >> 20) & 0x3ff);
r = (r << 6) | (r >> 4);
*(unsigned short *) &dst[dst_pos] = a;
*(unsigned short *) &dst[dst_pos + 2] = r;
*(unsigned short *) &dst[dst_pos + 4] = g;
*(unsigned short *) &dst[dst_pos + 6] = b;
}
}
}
#else
static const char ConverterUnpack_str[] =
"extern \"C\" {\n"
"__global__ void\n"
"GstHipConverterUnpack_RGB_RGBx\n"
"(unsigned char *src, unsigned char *dst, int width, int height,\n"
" int src_stride, int dst_stride)\n"
"{\n"
" int x_pos = blockIdx.x * blockDim.x + threadIdx.x;\n"
" int y_pos = blockIdx.y * blockDim.y + threadIdx.y;\n"
" if (x_pos < width && y_pos < height) {\n"
" int dst_pos = x_pos * 4 + y_pos * dst_stride;\n"
" int src_pos = x_pos * 3 + y_pos * src_stride;\n"
" dst[dst_pos] = src[src_pos];\n"
" dst[dst_pos + 1] = src[src_pos + 1];\n"
" dst[dst_pos + 2] = src[src_pos + 2];\n"
" dst[dst_pos + 3] = 0xff;\n"
" }\n"
"}\n"
"\n"
"__global__ void\n"
"GstHipConverterUnpack_RGB10A2_ARGB64\n"
"(unsigned char *src, unsigned char *dst, int width, int height,\n"
" int src_stride, int dst_stride)\n"
"{\n"
" int x_pos = blockIdx.x * blockDim.x + threadIdx.x;\n"
" int y_pos = blockIdx.y * blockDim.y + threadIdx.y;\n"
" if (x_pos < width && y_pos < height) {\n"
" unsigned short a, r, g, b;\n"
" unsigned int val;\n"
" int dst_pos = x_pos * 8 + y_pos * dst_stride;\n"
" val = *(unsigned int *)&src[x_pos * 4 + y_pos * src_stride];\n"
" a = (val >> 30) & 0x03;\n"
" a = (a << 14) | (a << 12) | (a << 10) | (a << 8) | (a << 6) | (a << 4) | (a << 2) | (a << 0);\n"
" r = (val & 0x3ff);\n"
" r = (r << 6) | (r >> 4);\n"
" g = ((val >> 10) & 0x3ff);\n"
" g = (g << 6) | (g >> 4);\n"
" b = ((val >> 20) & 0x3ff);\n"
" b = (b << 6) | (b >> 4);\n"
" *(unsigned short *) &dst[dst_pos] = a;\n"
" *(unsigned short *) &dst[dst_pos + 2] = r;\n"
" *(unsigned short *) &dst[dst_pos + 4] = g;\n"
" *(unsigned short *) &dst[dst_pos + 6] = b;\n"
" }\n"
"}\n"
"\n"
"__global__ void\n"
"GstHipConverterUnpack_BGR10A2_ARGB64\n"
"(unsigned char *src, unsigned char *dst, int width, int height,\n"
" int src_stride, int dst_stride)\n"
"{\n"
" int x_pos = blockIdx.x * blockDim.x + threadIdx.x;\n"
" int y_pos = blockIdx.y * blockDim.y + threadIdx.y;\n"
" if (x_pos < width && y_pos < height) {\n"
" unsigned short a, r, g, b;\n"
" unsigned int val;\n"
" int dst_pos = x_pos * 8 + y_pos * dst_stride;\n"
" val = *(unsigned int *)&src[x_pos * 4 + y_pos * src_stride];\n"
" a = (val >> 30) & 0x03;\n"
" a = (a << 14) | (a << 12) | (a << 10) | (a << 8) | (a << 6) | (a << 4) | (a << 2) | (a << 0);\n"
" b = (val & 0x3ff);\n"
" b = (b << 6) | (b >> 4);\n"
" g = ((val >> 10) & 0x3ff);\n"
" g = (g << 6) | (g >> 4);\n"
" r = ((val >> 20) & 0x3ff);\n"
" r = (r << 6) | (r >> 4);\n"
" *(unsigned short *) &dst[dst_pos] = a;\n"
" *(unsigned short *) &dst[dst_pos + 2] = r;\n"
" *(unsigned short *) &dst[dst_pos + 4] = g;\n"
" *(unsigned short *) &dst[dst_pos + 6] = b;\n"
" }\n"
"}\n"
"}\n"
"\n";
#endif

2935
subprojects/gst-plugins-bad/sys/hip/kernel/converter.cu Normal file
View File

File diff suppressed because it is too large Load Diff

89
subprojects/gst-plugins-bad/sys/hip/meson.build Normal file
View File

@ -0,0 +1,89 @@
hip_sources = [
'gsthipbasefilter.cpp',
'gsthipbufferpool.cpp',
'gsthipconverter.cpp',
'gsthipconvertscale.cpp',
'gsthipdevice.cpp',
'gsthipmemory.cpp',
'gsthipmemorycopy.cpp',
'gsthiprtc.cpp',
'gsthiputils.cpp',
'plugin.cpp',
]
doc_sources = []
foreach s: hip_sources
doc_sources += meson.current_source_dir() / s
endforeach
plugin_sources += {
'hip': pathsep.join(doc_sources)
}
extra_args = [
'-DGST_USE_UNSTABLE_API',
'-D__HIP_PLATFORM_AMD__',
'-D__HIP_DISABLE_CPP_FUNCTIONS__',
]
hip_option = get_option('hip')
if hip_option.disabled()
subdir_done()
endif
if host_system not in ['linux', 'windows']
subdir_done()
endif
hip_dep = dependency('', required: false)
hip_root = run_command(python3, '-c', 'import os; print(os.environ.get("HIP_PATH"))', check: false).stdout().strip()
if host_system == 'windows'
hip_root = run_command(python3, '-c', 'import os; print(os.environ.get("HIP_PATH"))', check: false).stdout().strip()
else
# HIP does not offer pkg-config it seems
hip_root = '/opt/rocm'
endif
if hip_root != '' and hip_root != 'None'
hip_lib_dir = join_paths(hip_root, 'lib')
hip_lib = cc.find_library('amdhip64', dirs: hip_lib_dir,
required : hip_option)
hiprtc_lib = cc.find_library('hiprtc', dirs: hip_lib_dir,
required : hip_option)
hip_inc_dir = include_directories(join_paths(hip_root, 'include'))
has_hip_header = cc.has_header('hip/hip_runtime.h',
include_directories: hip_inc_dir,
required: hip_option)
has_hiprtc_header = cc.has_header('hip/hiprtc.h',
include_directories: hip_inc_dir,
required: hip_option)
if hip_lib.found() and hiprtc_lib.found() and has_hip_header and has_hiprtc_header
hip_dep = declare_dependency(include_directories: hip_inc_dir,
dependencies: [hip_lib, hiprtc_lib])
endif
endif
if not hip_dep.found()
if hip_option.enabled()
error('The hip was enabled explicitly, but required dependencies were not found')
endif
subdir_done()
endif
extra_args += cc.get_supported_arguments([
'-Wno-undef'
])
gsthip = library('gsthip', hip_sources,
c_args : gst_plugins_bad_args + extra_args,
cpp_args: gst_plugins_bad_args + extra_args,
include_directories : [configinc],
dependencies : [gstbase_dep, gstvideo_dep, hip_dep],
install : true,
install_dir : plugins_install_dir,
)
plugins += [gsthip]

64
subprojects/gst-plugins-bad/sys/hip/plugin.cpp Normal file
View File

@ -0,0 +1,64 @@
/* GStreamer
* Copyright (C) 2025 Seungha Yang <seungha@centricular.com>
*
* This library is free software; you can redistribute it and/or
* modify it under the terms of the GNU Library General Public
* License as published by the Free Software Foundation; either
* version 2 of the License, or (at your option) any later version.
*
* This library 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
* Library General Public License for more details.
*
* You should have received a copy of the GNU Library General Public
* License along with this library; if not, write to the
* Free Software Foundation, Inc., 51 Franklin St, Fifth Floor,
* Boston, MA 02110-1301, USA.
*/
#ifdef HAVE_CONFIG_H
#include <config.h>
#endif
#include <gst/gst.h>
#include "gsthipdevice.h"
#include "gsthipmemorycopy.h"
#include "gsthipconvertscale.h"
static gboolean
plugin_init (GstPlugin * plugin)
{
auto device = gst_hip_device_new (0);
if (!device)
return TRUE;
gst_element_register (plugin,
"hipupload", GST_RANK_NONE, GST_TYPE_HIP_UPLOAD);
gst_element_register (plugin,
"hipdownload", GST_RANK_NONE, GST_TYPE_HIP_DOWNLOAD);
gboolean texture_support = FALSE;
g_object_get (device, "texture2d-support", &texture_support, nullptr);
if (texture_support) {
gst_element_register (plugin,
"hipconvertscale", GST_RANK_NONE, GST_TYPE_HIP_CONVERT_SCALE);
gst_element_register (plugin,
"hipconvert", GST_RANK_NONE, GST_TYPE_HIP_CONVERT);
gst_element_register (plugin,
"hipscale", GST_RANK_NONE, GST_TYPE_HIP_SCALE);
} else {
gst_plugin_add_status_info (plugin,
"Texture2D not supported by HIP device");
}
gst_clear_object (&device);
return TRUE;
}
GST_PLUGIN_DEFINE (GST_VERSION_MAJOR,
GST_VERSION_MINOR,
hip,
"HIP plugin",
plugin_init, VERSION, "LGPL", GST_PACKAGE_NAME, GST_PACKAGE_ORIGIN)

1
subprojects/gst-plugins-bad/sys/meson.build
View File

@ -13,6 +13,7 @@ subdir('directshow')
subdir('dvb')
subdir('dwrite')
subdir('fbdev')
subdir('hip')
subdir('ipcpipeline')
subdir('kms')
subdir('magicleap')