/*
Copyright (c) 2015 - 2022 Advanced Micro Devices, Inc. All rights reserved.
Permission is hereby granted, free of charge, to any person obtaining a copy
of this software and associated documentation files (the "Software"), to deal
in the Software without restriction, including without limitation the rights
to use, copy, modify, merge, publish, distribute, sublicense, and/or sell
copies of the Software, and to permit persons to whom the Software is
furnished to do so, subject to the following conditions:
The above copyright notice and this permission notice shall be included in
all copies or substantial portions of the Software.
THE SOFTWARE IS PROVIDED "AS IS", WITHOUT WARRANTY OF ANY KIND, EXPRESS OR
IMPLIED, INCLUDING BUT NOT LIMITED TO THE WARRANTIES OF MERCHANTABILITY,
FITNESS FOR A PARTICULAR PURPOSE AND NONINFRINGEMENT. IN NO EVENT SHALL THE
AUTHORS OR COPYRIGHT HOLDERS BE LIABLE FOR ANY CLAIM, DAMAGES OR OTHER
LIABILITY, WHETHER IN AN ACTION OF CONTRACT, TORT OR OTHERWISE, ARISING FROM,
OUT OF OR IN CONNECTION WITH THE SOFTWARE OR THE USE OR OTHER DEALINGS IN
THE SOFTWARE.
*/
#include "ago_internal.h"
#define VX_MAX_TENSOR_DIMENSIONS 6
static inline vx_uint32 vxComputePatchOffset(vx_uint32 x, vx_uint32 y, const vx_imagepatch_addressing_t *addr)
{
#if VX_SCALE_UNITY == (1024u)
return ((addr->stride_y * ((addr->scale_y * y) >> 10)) +
(addr->stride_x * ((addr->scale_x * x) >> 10)));
#else
return ((addr->stride_y * ((addr->scale_y * y) / VX_SCALE_UNITY)) +
(addr->stride_x * ((addr->scale_x * x) / VX_SCALE_UNITY)));
#endif
}
/*! \brief Creates a \ref vx_context.
* \details This creates a top-level object context for OpenVX.
* \note This is required to do anything else.
* \returns The reference to the implementation context.
* \retval 0 No context was created.
* \retval * A context reference.
* \ingroup group_context
* \post \ref vxReleaseContext
*/
VX_API_ENTRY vx_context VX_API_CALL vxCreateContext()
{
vx_context context = agoCreateContextFromPlatform(nullptr);
return context;
}
/*! \brief Releases the OpenVX object context.
* \details All reference counted objects are garbage-collected by the return of this call.
* No calls are possible using the parameter context after the context has been
* released until a new reference from \ref vxCreateContext is returned.
* All outstanding references to OpenVX objects from this context are invalid
* after this call.
* \param [in] context The pointer to the reference to the context.
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
* \ingroup group_context
* \pre \ref vxCreateContext
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseContext(vx_context *context)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (context && !agoReleaseContext(*context)) {
*context = NULL;
status = VX_SUCCESS;
}
return status;
}
/**
* \brief Set custom image format description.
* \ingroup vx_framework_reference
*
* This function is used to support custom image formats with single-plane by ISVs. Should be called from vxPublishKernels().
*
* \param [in] context The context.
* \param [in] format The image format.
* \param [in] desc The image format description.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE if reference is not valid.
* \retval VX_ERROR_INVALID_FORMAT if format is already in use.
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetContextImageFormatDescription(vx_context context, vx_df_image format, const AgoImageFormatDescription * desc)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context)) {
status = VX_ERROR_INVALID_FORMAT;
if (desc->planes == 1 && !agoSetImageComponentsAndPlanes(context, format, desc->components, desc->planes, (vx_uint32)desc->pixelSizeInBitsNum, (vx_uint32)(desc->pixelSizeInBitsDenom ? desc->pixelSizeInBitsDenom : 1), desc->colorSpace, desc->channelRange)) {
status = VX_SUCCESS;
}
}
return status;
}
/**
* \brief Get custom image format description.
* \ingroup vx_framework_reference
* \param [in] context The context.
* \param [in] format The image format.
* \param [out] desc The image format description.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE if reference is not valid.
* \retval VX_ERROR_INVALID_FORMAT if format is already in use.
*/
VX_API_ENTRY vx_status VX_API_CALL vxGetContextImageFormatDescription(vx_context context, vx_df_image format, AgoImageFormatDescription * desc)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context)) {
status = VX_ERROR_INVALID_FORMAT;
vx_uint32 pixelSizeInBitsNum, pixelSizeInBitsDenom;
if (!agoGetImageComponentsAndPlanes(context, format, &desc->components, &desc->planes, &pixelSizeInBitsNum, &pixelSizeInBitsDenom, &desc->colorSpace, &desc->channelRange)) {
desc->pixelSizeInBitsNum = pixelSizeInBitsNum;
desc->pixelSizeInBitsDenom = pixelSizeInBitsDenom;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Retrieves the context from any reference from within a context.
* \param [in] reference The reference from which to extract the context.
* \ingroup group_context
* \return The overall context that created the particular
* reference.
*/
VX_API_ENTRY vx_context VX_API_CALL vxGetContext(vx_reference reference)
{
vx_context context = NULL;
if (agoIsValidReference(reference)) {
context = reference->context;
}
return context;
}
/*! \brief Queries the context for some specific information.
* \param [in] context The reference to the context.
* \param [in] attribute The attribute to query. Use a \ref vx_context_attribute_e.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the context is not a \ref vx_context.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \retval VX_ERROR_NOT_SUPPORTED If the attribute is not supported on this implementation.
* \ingroup group_context
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryContext(vx_context context, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
CAgoLock lock(context->cs);
switch (attribute)
{
case VX_CONTEXT_ATTRIBUTE_VENDOR_ID:
if (size == sizeof(vx_uint16)) {
*(vx_uint16 *)ptr = VX_ID_AMD;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_VERSION:
if (size == sizeof(vx_uint16)) {
*(vx_uint16 *)ptr = (vx_uint16)VX_VERSION;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_MODULES:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = (vx_uint32)context->num_active_modules;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_REFERENCES:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = (vx_uint32)context->num_active_references;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_IMPLEMENTATION:
if (size <= VX_MAX_IMPLEMENTATION_NAME) {
strncpy((char *)ptr, "AMD OpenVX " AGO_VERSION, VX_MAX_IMPLEMENTATION_NAME);
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_EXTENSIONS_SIZE:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = strlen(context->extensions) + 1;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_EXTENSIONS:
if (size >= strlen(context->extensions) + 1) {
strcpy((char *)ptr, context->extensions);
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_CONVOLUTION_MAXIMUM_DIMENSION:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = AGO_MAX_CONVOLUTION_DIM;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_NONLINEAR_MAX_DIMENSION:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = AGO_MAX_NONLINEAR_FILTER_DIM;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_OPTICAL_FLOW_WINDOW_MAXIMUM_DIMENSION:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = AGO_OPTICALFLOWPYRLK_MAX_DIM;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_IMMEDIATE_BORDER_MODE:
if (size == sizeof(vx_border_mode_t)) {
*(vx_border_mode_t *)ptr = context->immediate_border_mode;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_IMMEDIATE_BORDER_POLICY:
if (size == sizeof(vx_border_mode_policy_e)) {
*(vx_border_mode_policy_e *)ptr = context->immediate_border_policy;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_UNIQUE_KERNELS:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = (vx_uint32)context->kernelList.count;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_UNIQUE_KERNEL_TABLE:
if (size == (context->kernelList.count * sizeof(vx_kernel_info_t))) {
vx_kernel_info_t * table = (vx_kernel_info_t *)ptr;
for (AgoKernel * kernel = context->kernelList.head; kernel; kernel = kernel->next, table++) {
table->enumeration = kernel->id;
strncpy(table->name, kernel->name, VX_MAX_KERNEL_NAME);
}
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_ATTRIBUTE_AMD_AFFINITY:
if (size == sizeof(AgoTargetAffinityInfo_)) {
*(AgoTargetAffinityInfo_ *)ptr = context->attr_affinity;
status = VX_SUCCESS;
}
break;
#if ENABLE_OPENCL
case VX_CONTEXT_ATTRIBUTE_AMD_OPENCL_CONTEXT:
if (size == sizeof(cl_context)) {
if (!context->opencl_context && agoGpuOclCreateContext(context, nullptr) != VX_SUCCESS) {
status = VX_FAILURE;
}
else {
*(cl_context *)ptr = context->opencl_context;
status = VX_SUCCESS;
}
}
else if (size == 0) {
// special case to just request internal context without creating one
// when not available
*(cl_context *)ptr = context->opencl_context;
status = VX_SUCCESS;
}
break;
case VX_CONTEXT_CL_QUEUE_PROPERTIES:
if (size == sizeof(cl_command_queue_properties)) {
*(cl_command_queue_properties *)ptr = context->opencl_cmdq_properties;
status = VX_SUCCESS;
}
break;
#elif ENABLE_HIP
case VX_CONTEXT_ATTRIBUTE_AMD_HIP_DEVICE:
if (size == sizeof(int)) {
if (context->hip_device_id < 0 && agoGpuHipCreateContext(context, -1) != VX_SUCCESS) {
status = VX_FAILURE;
}
else {
*(int *)ptr = context->hip_device_id;
status = VX_SUCCESS;
}
}
else if (size == 0) {
// special case to just request internal context without creating one
// when not available
*(int *)ptr = context->hip_device_id;
status = VX_SUCCESS;
}
break;
#endif
case VX_CONTEXT_MAX_TENSOR_DIMS:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = AGO_MAX_TENSOR_DIMENSIONS;
status = VX_SUCCESS;
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Sets an attribute on the context.
* \param [in] context The handle to the overall context.
* \param [in] attribute The attribute to set from \ref vx_context_attribute_e.
* \param [in] ptr The pointer to the data to which to set the attribute.
* \param [in] size The size in bytes of the data to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the context is not a \ref vx_context.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \retval VX_ERROR_NOT_SUPPORTED If the attribute is not settable.
* \ingroup group_context
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetContextAttribute(vx_context context, vx_enum attribute, const void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context)) {
status = VX_ERROR_INVALID_PARAMETERS;
CAgoLock lock(context->cs);
switch (attribute)
{
case VX_CONTEXT_ATTRIBUTE_IMMEDIATE_BORDER_MODE:
if(!ptr) return VX_ERROR_INVALID_PARAMETERS;
if (size == sizeof(vx_border_mode_t)) {
vx_border_mode_t immediate_border_mode = *(vx_border_mode_t *)ptr;
if (immediate_border_mode.mode == VX_BORDER_MODE_UNDEFINED || immediate_border_mode.mode == VX_BORDER_MODE_CONSTANT || immediate_border_mode.mode == VX_BORDER_MODE_REPLICATE) {
context->immediate_border_mode = immediate_border_mode;
if (immediate_border_mode.mode == VX_BORDER_MODE_UNDEFINED || immediate_border_mode.mode == VX_BORDER_MODE_REPLICATE)
memset(&context->immediate_border_mode.constant_value, 0, sizeof(context->immediate_border_mode.constant_value));
status = VX_SUCCESS;
}
else
status = VX_ERROR_INVALID_VALUE;
}
break;
case VX_CONTEXT_ATTRIBUTE_AMD_SET_TEXT_MACRO:
if(!ptr) return VX_ERROR_INVALID_PARAMETERS;
if (size == sizeof(AgoContextTextMacroInfo)) {
status = VX_SUCCESS;
AgoContextTextMacroInfo * info = (AgoContextTextMacroInfo *)ptr;
for (auto it = context->macros.begin(); it != context->macros.end(); ++it) {
if (!strcmp(it->name, info->macroName)) {
status = VX_FAILURE;
agoAddLogEntry(&context->ref, status, "ERROR: vxSetContextAttribute: macro already exists: %s\n", info->macroName);
break;
}
}
if (status == VX_SUCCESS) {
MacroData macro;
macro.text = macro.text_allocated = (char *)calloc(1, strlen(info->text) + 1);
if (!macro.text) {
status = VX_ERROR_NO_MEMORY;
}
else {
strncpy(macro.name, info->macroName, sizeof(macro.name) - 1);
strcpy(macro.text, info->text);
context->macros.push_back(macro);
}
}
}
break;
case VX_CONTEXT_ATTRIBUTE_AMD_SET_MERGE_RULE:
if(!ptr) return VX_ERROR_INVALID_PARAMETERS;
if (size == sizeof(AgoNodeMergeRule)) {
status = VX_SUCCESS;
context->merge_rules.push_back(*(AgoNodeMergeRule *)ptr);
}
break;
case VX_CONTEXT_ATTRIBUTE_AMD_AFFINITY:
if(!ptr) return VX_ERROR_INVALID_PARAMETERS;
if (size == sizeof(AgoTargetAffinityInfo_)) {
status = VX_SUCCESS;
context->attr_affinity = *(AgoTargetAffinityInfo_ *)ptr;
}
break;
#if ENABLE_OPENCL
case VX_CONTEXT_ATTRIBUTE_AMD_OPENCL_CONTEXT:
if(!ptr) return VX_ERROR_INVALID_PARAMETERS;
if (size == sizeof(cl_context)) {
if (!context->opencl_context) {
status = agoGpuOclCreateContext(context, *(cl_context *)ptr);
}
else {
status = VX_FAILURE;
}
}
break;
case VX_CONTEXT_CL_QUEUE_PROPERTIES:
if(!ptr) return VX_ERROR_INVALID_PARAMETERS;
if (size == sizeof(cl_command_queue_properties)) {
context->opencl_cmdq_properties = *(cl_command_queue_properties *)ptr;
status = VX_SUCCESS;
}
break;
#elif ENABLE_HIP
case VX_CONTEXT_ATTRIBUTE_AMD_HIP_DEVICE:
if (size == sizeof(hipDevice_t)) {
if (*(int *)ptr >= 0 && agoGpuHipCreateContext(context, *(int *)ptr) == VX_SUCCESS) {
status = VX_SUCCESS;
}
else {
status = VX_FAILURE;
}
}
break;
#endif
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
return status;
}
/*! \brief Provides a generic API to give platform-specific hints to the implementation.
* \param [in] reference The reference to the object to hint at.
* This could be \ref vx_context, \ref vx_graph, \ref vx_node, \ref vx_image, \ref vx_array, or any other reference.
* \param [in] hint A \ref vx_hint_e \a hint to give to a \ref vx_context. This is a platform-specific optimization or implementation mechanism.
* \param [in] data Optional vendor specific data.
* \param [in] data_size Size of the data structure \p data.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No error.
* \retval VX_ERROR_INVALID_REFERENCE If context or reference is invalid.
* \retval VX_ERROR_NOT_SUPPORTED If the hint is not supported.
* \ingroup group_hint
*/
VX_API_ENTRY vx_status VX_API_CALL vxHint(vx_reference reference, vx_enum hint, const void* data, vx_size data_size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidReference(reference)) {
vx_context context = reference->context;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
status = VX_SUCCESS;
switch (hint)
{
case VX_HINT_SERIALIZE:
reference->hint_serialize = true;
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Provides a generic API to give platform-specific directives to the implementations.
* \param [in] context The reference to the implementation context.
* \param [in] reference The reference to the object to set the directive on.
* This could be \ref vx_context, \ref vx_graph, \ref vx_node, \ref vx_image, \ref vx_array, or any other reference.
* \param [in] directive The directive to set.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No error.
* \retval VX_ERROR_INVALID_REFERENCE If context or reference is invalid.
* \retval VX_ERROR_NOT_SUPPORTED If the directive is not supported.
* \ingroup group_directive
*/
VX_API_ENTRY vx_status VX_API_CALL vxDirective(vx_reference reference, vx_enum directive)
{
return agoDirective(reference, directive);
}
/*! \brief Provides a generic API to return status values from Object constructors if they
* fail.
* \note Users do not need to strictly check every object creator as the errors
* should properly propogate and be detected during verification time or run-time.
* \code
* vx_image img = vxCreateImage(context, 639, 480, VX_DF_IMAGE_UYVY);
* vx_status status = vxGetStatus((vx_reference)img);
* // status == VX_ERROR_INVALID_DIMENSIONS
* vxReleaseImage(&img);
* \endcode
* \pre Appropriate Object Creator function.
* \post Appropriate Object Release function.
* \param [in] reference The reference to check for construction errors.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No error.
* \retval * Some error occurred, please check enumeration list and constructor.
* \ingroup group_basic_features
*/
VX_API_ENTRY vx_status VX_API_CALL vxGetStatus(vx_reference reference)
{
if(reference == NULL)
return VX_ERROR_NO_RESOURCES;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidReference(reference)) {
status = reference->status;
}
return status;
}
/*!
* \brief Registers user-defined structures to the context.
* \param [in] context The reference to the implementation context.
* \param [in] size The size of user struct in bytes.
* \return A \ref vx_enum value that is a type given to the User
* to refer to their custom structure when declaring a \ref vx_array
* of that structure.
* \retval VX_TYPE_INVALID If the namespace of types has been exhausted.
* \note This call should only be used once within the lifetime of a context for
* a specific structure.
*
* \snippet vx_arrayrange.c array define
* \ingroup group_adv_array
*/
VX_API_ENTRY vx_enum VX_API_CALL vxRegisterUserStruct(vx_context context, vx_size size)
{
vx_enum type = VX_TYPE_INVALID;
if (agoIsValidContext(context) && (size > 0)) {
CAgoLock lock(context->cs);
type = agoAddUserStruct(context, size, NULL);
}
return type;
}
/*!
* \brief Registers user-defined structures to the context, and associates a name to it.
* \param [in] context The reference to the implementation context.
* \param [in] size The size of user struct in bytes.
* \param [in] *type_name Pointer to the '\0' terminated string that identifies the
* user struct type. The string is copied by the function so
* that it stays the property of the caller. NULL means that
* the user struct is not named. The length of the string
* shall be lower than VX_MAX_REFERENCE_NAME bytes.
* \return A \ref vx_enum value that is a type given to the User
* to refer to their custom structure when declaring a \ref vx_array
* of that structure.
* \retval VX_TYPE_INVALID If the namespace of types has been exhausted.
* \note This call should only be used once within the lifetime of a context for
* a specific structure.
* \ingroup group_adv_array
*/
VX_API_ENTRY vx_enum VX_API_CALL vxRegisterUserStructWithName(vx_context context, vx_size size, const vx_char* type_name)
{
vx_enum type = VX_TYPE_INVALID;
if (agoIsValidContext(context) && (size > 0)) {
CAgoLock lock(context->cs);
if((sizeof(char)*strlen(type_name)) < VX_MAX_REFERENCE_NAME || type_name == NULL)
type = agoAddUserStruct(context, size, (vx_char*)type_name);
}
return type;
}
/*!
* \brief Returns the name of the user-defined structure associated with the enumeration given.
* \param [in] context The reference to the implementation context.
* \param [in] type_name The enumeration value of the user struct
* \param [out] name_size Name of the user struct
* \param [in] name_size The size of allocated buffer pointed to by type_name
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS user_struct_type was valid, and name was found and returned
* \retval VX_ERROR_INVALID_PARAMETERS user_struct_type was not a valid user struct enumeration.
* \retval VX_ERROR_NO_MEMORY name_size is too small to hold the name of the user struct type.
* \retval VX_FAILURE user_struct_type does not have an associated type name.
* \pre \ref vxRegisterUserStructWithName should be called for this user struct.
* \ingroup group_adv_array
*/
VX_API_ENTRY vx_status VX_API_CALL vxGetUserStructNameByEnum(vx_context context, vx_enum user_struct_type, vx_char* type_name, vx_size name_size)
{
vx_status status = VX_ERROR_INVALID_PARAMETERS;
if (agoIsValidContext(context) && (user_struct_type!= VX_TYPE_INVALID)) {
status = VX_FAILURE;
const char* temp_type_name = (agoGetUserStructName(context, user_struct_type));
if(temp_type_name != NULL) {
status = VX_ERROR_NO_MEMORY;
if((strlen(temp_type_name) + 1) <= name_size) {
memcpy(type_name, temp_type_name, strlen(temp_type_name) + 1);
status = VX_SUCCESS;
}
}
}
return status;
}
/*!
* \brief Returns the enum of the user-defined structure associated with the name given
* \param [in] context The reference to the implementation context.
* \param [in] type_name Pointer to the '\0' terminated string that identifies the user
* struct type. The length of the string shall be lower than VX_MAX_REFERENCE_NAME bytes.
* \param [out] user_struct_type The enumeration value of the user struct
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS type_name was valid, and enumeration was found and returned
* \retval VX_FAILURE type_name does not match any user struct enumeration.
* \pre \ref vxRegisterUserStructWithName should be called for this user struct.
* \ingroup group_adv_array
*/
VX_API_ENTRY vx_status VX_API_CALL vxGetUserStructEnumByName(vx_context context, const vx_char* type_name, vx_enum *user_struct_type)
{
vx_status status = VX_FAILURE;
if (agoIsValidContext(context) && type_name) {
if((strlen(type_name)) < VX_MAX_REFERENCE_NAME) {
*user_struct_type = agoGetUserStructType(context, (vx_char*)type_name);
if(*user_struct_type != 0){
status = VX_SUCCESS;
}
}
}
return status;
}
/*!
* \brief Allocates and registers user-defined kernel enumeration to a context.
* The allocated enumeration is from available pool of 4096 enumerations reserved
* for dynamic allocation from VX_KERNEL_BASE(VX_ID_USER,0).
* \param [in] context The reference to the implementation context.
* \param [out] pKernelEnumId pointer to return \ref vx_enum for user-defined kernel.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_NO_RESOURCES The enumerations has been exhausted.
* \ingroup group_user_kernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxAllocateUserKernelId(vx_context context, vx_enum * pKernelEnumId)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context) && pKernelEnumId)
{
status = VX_ERROR_NO_RESOURCES;
if (context->nextUserKernelId <= VX_KERNEL_MASK)
{
*pKernelEnumId = VX_KERNEL_BASE(VX_ID_USER, 0) + context->nextUserKernelId++;
status = VX_SUCCESS;
}
}
return status;
}
/*!
* \brief Allocates and registers user-defined kernel library ID to a context.
*
* The allocated library ID is from available pool of library IDs (1..255)
* reserved for dynamic allocation. The returned libraryId can be used by
* user-kernel library developer to specify individual kernel enum IDs in
* a header file, shown below:
* \code
* #define MY_KERNEL_ID1(libraryId) (VX_KERNEL_BASE(VX_ID_USER,libraryId) + 0);
* #define MY_KERNEL_ID2(libraryId) (VX_KERNEL_BASE(VX_ID_USER,libraryId) + 1);
* #define MY_KERNEL_ID3(libraryId) (VX_KERNEL_BASE(VX_ID_USER,libraryId) + 2);
* \endcode
* \param [in] context The reference to the implementation context.
* \param [out] pLibraryId pointer to \ref vx_enum for user-kernel libraryId.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_NO_RESOURCES The enumerations has been exhausted.
* \ingroup group_user_kernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxAllocateUserKernelLibraryId(vx_context context, vx_enum * pLibraryId)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context) && pLibraryId)
{
status = VX_ERROR_NO_RESOURCES;
if (context->nextUserLibraryId <= VX_LIBRARY(VX_LIBRARY_MASK))
{
*pLibraryId = context->nextUserLibraryId++;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Sets the default target of the immediate mode. Upon successful execution of this
* function any future execution of immediate mode function is attempted on the new default
* target of the context.
* \param [in] context The reference to the implementation context.
* \param [in] target_enum The default immediate mode target enum to be set
* to the \ref vx_context object. Use a \ref vx_target_e.
* \param [in] target_string The target name ASCII string. This contains a valid value
* when target_enum is set to \ref VX_TARGET_STRING, otherwise it is ignored.
* \ingroup group_context
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Default target set.
* \retval VX_ERROR_INVALID_REFERENCE If the context is not a \ref vx_context.
* \retval VX_ERROR_NOT_SUPPORTED If the specified target is not supported in this context.
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetImmediateModeTarget(vx_context context, vx_enum target_enum, const char* target_string)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context))
{
status = VX_ERROR_NOT_SUPPORTED;
if (target_enum == VX_TARGET_ANY) {
status = VX_SUCCESS;
}
else if (target_enum == VX_TARGET_STRING) {
if (!target_string) {
status = VX_ERROR_INVALID_REFERENCE;
}
else if (!_stricmp(target_string, "any") || !_stricmp(target_string, "cpu")) {
// only cpu mode is supported as immediate mode target
status = VX_SUCCESS;
}
}
}
return status;
}
/*==============================================================================
IMAGE
=============================================================================*/
/*! \brief Creates an opaque reference to an image buffer.
* \details Not guaranteed to exist until the \ref vx_graph containing it has been verified.
* \param [in] context The reference to the implementation context.
* \param [in] width The image width in pixels.
* \param [in] height The image height in pixels.
* \param [in] color The VX_DF_IMAGE (\ref vx_df_image_e) code that represents the format of the image and the color space.
* \return An image reference or zero when an error is encountered.
* \see vxAccessImagePatch to obtain direct memory access to the image data.
* \ingroup group_image
*/
VX_API_ENTRY vx_image VX_API_CALL vxCreateImage(vx_context context, vx_uint32 width, vx_uint32 height, vx_df_image color)
{
AgoData * data = NULL;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
char desc[128]; sprintf(desc, "image:%4.4s,%d,%d", FORMAT_STR(color), width, height);
data = agoCreateDataFromDescription(context, NULL, desc, true);
if (data) {
agoGenerateDataName(context, "image", data->name);
agoAddData(&context->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&context->dataList, data->children[i]);
}
}
}
}
return (vx_image)data;
}
/*! \brief Creates an image from another image given a rectangle. This second
* reference refers to the data in the original image. Updates to this image
* updates the parent image. The rectangle must be defined within the pixel space
* of the parent image.
* \param [in] img The reference to the parent image.
* \param [in] rect The region of interest rectangle. Must contain points within
* the parent image pixel space.
* \return The reference to the sub-image or zero if the rectangle is invalid.
* \ingroup group_image
*/
VX_API_ENTRY vx_image VX_API_CALL vxCreateImageFromROI(vx_image img, const vx_rectangle_t *rect)
{
AgoData * master_img = (AgoData *)img;
AgoData * data = NULL;
if (agoIsValidData(master_img, VX_TYPE_IMAGE)) {
vx_context context = master_img->ref.context;
CAgoLock lock(context->cs);
char desc[128]; sprintf(desc, "image-roi:%s,%d,%d,%d,%d", master_img->name.c_str(), rect->start_x, rect->start_y, rect->end_x, rect->end_y);
data = agoCreateDataFromDescription(context, NULL, desc, true);
if (data) {
agoGenerateDataName(context, "image-roi", data->name);
agoAddData(&context->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&context->dataList, data->children[i]);
}
}
}
}
return (vx_image)data;
}
/*! \brief Creates a reference to an image object that has a singular,
* uniform value in all pixels.
* \details The value pointer must reflect the specific format of the desired
* image. For example:
* | Color | Value Ptr |
* |:------------|:-----------|
* | \ref VX_DF_IMAGE_U8 | vx_uint8 * |
* | \ref VX_DF_IMAGE_S16 | vx_int16 * |
* | \ref VX_DF_IMAGE_U16 | vx_uint16 *|
* | \ref VX_DF_IMAGE_S32 | vx_int32 * |
* | \ref VX_DF_IMAGE_U32 | vx_uint32 *|
* | \ref VX_DF_IMAGE_RGB | vx_uint8 pixel[3] in R, G, B order |
* | \ref VX_DF_IMAGE_RGBX | vx_uint8 pixels[4] |
* | Any YUV | vx_uint8 pixel[3] in Y, U, V order |
*
* \param [in] context The reference to the implementation context.
* \param [in] width The image width in pixels.
* \param [in] height The image height in pixels.
* \param [in] color The VX_DF_IMAGE (\ref vx_df_image_e) code that represents the format of the image and the color space.
* \param [in] value The pointer to the pixel value to which to set all pixels.
* \return An image reference or zero when an error is encountered.
* \see vxAccessImagePatch to obtain direct memory access to the image data.
* \note \ref vxAccessImagePatch and \ref vxCommitImagePatch may be called with
* a uniform image reference.
* \ingroup group_image
*/
VX_API_ENTRY vx_image VX_API_CALL vxCreateUniformImage(vx_context context, vx_uint32 width, vx_uint32 height, vx_df_image color, const vx_pixel_value_t *value)
{
AgoData * data = NULL;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
char desc[128];
if (color == VX_DF_IMAGE_U8) {
sprintf(desc, "image-uniform:%4.4s,%d,%d,%d", FORMAT_STR(color), width, height, value->U8);
}
else if (color == VX_DF_IMAGE_S16) {
sprintf(desc, "image-uniform:%4.4s,%d,%d,%d", FORMAT_STR(color), width, height, value->S16);
}
else if (color == VX_DF_IMAGE_U16) {
sprintf(desc, "image-uniform:%4.4s,%d,%d,%d", FORMAT_STR(color), width, height, value->U16);
}
else if (color == VX_DF_IMAGE_S32) {
sprintf(desc, "image-uniform:%4.4s,%d,%d,%d", FORMAT_STR(color), width, height, value->S32);
}
else if (color == VX_DF_IMAGE_U32) {
sprintf(desc, "image-uniform:%4.4s,%d,%d,%u", FORMAT_STR(color), width, height, value->U32);
}
else {
sprintf(desc, "image-uniform:%4.4s,%d,%d,%d,%d,%d,%d", FORMAT_STR(color), width, height, value->reserved[0], value->reserved[1], value->reserved[2], value->reserved[3]);
}
data = agoCreateDataFromDescription(context, NULL, desc, true);
if (data) {
agoGenerateDataName(context, "image-uniform", data->name);
agoAddData(&context->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&context->dataList, data->children[i]);
}
}
}
}
return (vx_image)data;
}
/*! \brief Creates an opaque reference to an image buffer with no direct
* user access. This function allows setting the image width, height, or format.
* \details Virtual data objects allow users to connect various nodes within a
* graph via data references without access to that data, but they also permit the
* implementation to take maximum advantage of possible optimizations. Use this
* API to create a data reference to link two or more nodes together when the
* intermediate data are not required to be accessed by outside entities. This API
* in particular allows the user to define the image format of the data without
* requiring the exact dimensions. Virtual objects are scoped within the graph
* they are declared a part of, and can't be shared outside of this scope.
* All of the following constructions of virtual images are valid.
* \code
* vx_context context = vxCreateContext();
* vx_graph graph = vxCreateGraph(context);
* vx_image virt[] = {
* vxCreateVirtualImage(graph, 0, 0, VX_DF_IMAGE_U8), // no specified dimension
* vxCreateVirtualImage(graph, 320, 240, VX_DF_IMAGE_VIRT), // no specified format
* vxCreateVirtualImage(graph, 640, 480, VX_DF_IMAGE_U8), // no user access
* };
* \endcode
* \param [in] graph The reference to the parent graph.
* \param [in] width The width of the image in pixels. A value of zero informs the interface that the value is unspecified.
* \param [in] height The height of the image in pixels. A value of zero informs the interface that the value is unspecified.
* \param [in] color The VX_DF_IMAGE (\ref vx_df_image_e) code that represents the format of the image and the color space. A value of \ref VX_DF_IMAGE_VIRT informs the interface that the format is unspecified.
* \return An image reference or zero when an error is encountered.
* \note Passing this reference to \ref vxAccessImagePatch will return an error.
* \ingroup group_image
*/
VX_API_ENTRY vx_image VX_API_CALL vxCreateVirtualImage(vx_graph graph, vx_uint32 width, vx_uint32 height, vx_df_image color)
{
AgoData * data = NULL;
if (agoIsValidGraph(graph)) {
vx_context context = graph->ref.context;
CAgoLock lock(graph->cs);
char desc[128]; sprintf(desc, "image-virtual:%4.4s,%d,%d", FORMAT_STR(color), width, height);
data = agoCreateDataFromDescription(context, graph, desc, true);
if (data) {
agoGenerateVirtualDataName(graph, "image", data->name);
agoAddData(&graph->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&graph->dataList, data->children[i]);
}
}
}
}
return (vx_image)data;
}
/*! \brief Creates a reference to an image object that was externally allocated.
* \param [in] context The reference to the implementation context.
* \param [in] color See the \ref vx_df_image_e codes. This mandates the
* number of planes needed to be valid in the \a addrs and \a ptrs arrays based on the format given.
* \param [in] addrs[] The array of image patch addressing structures that
* define the dimension and stride of the array of pointers. See note below.
* \param [in] ptrs[] The array of platform-defined references to each plane. See note below.
* \param [in] memory_type \ref vx_memory_type_e. When giving \ref VX_MEMORY_TYPE_HOST
* the \a ptrs array is assumed to be HOST accessible pointers to memory.
* \returns An image reference \ref vx_image. Any possible errors preventing a
* successful creation should be checked using \ref vxGetStatus.
* \note The user must call vxMapImagePatch prior to accessing the pixels of an image, even if the
* image was created via \ref vxCreateImageFromHandle. Reads or writes to memory referenced
* by ptrs[ ] after calling \ref vxCreateImageFromHandle without first calling
* \ref vxMapImagePatch will result in undefined behavior.
* The property of addr[] and ptrs[] arrays is kept by the caller (It means that the implementation will
* make an internal copy of the provided information. \a addr and \a ptrs can then simply be application's
* local variables).
* Only \a dim_x, \a dim_y, \a stride_x and \a stride_y fields of the \ref vx_imagepatch_addressing_t need to be
* provided by the application. Other fields (\a step_x, \a step_y, \a scale_x & \a scale_y) are ignored by this function.
* The layout of the imported memory must follow a row-major order. In other words, \a stride_x should be
* sufficiently large so that there is no overlap between data elements corresponding to different
* pixels, and \a stride_y >= \a stride_x * \a dim_x.
*
* In order to release the image back to the application we should use \ref vxSwapImageHandle.
*
* Import type of the created image is available via the image attribute \ref vx_image_attribute_e parameter.
*
* \ingroup group_image
*/
VX_API_ENTRY vx_image VX_API_CALL vxCreateImageFromHandle(vx_context context, vx_df_image color, const vx_imagepatch_addressing_t addrs[], void *const ptrs[], vx_enum memory_type)
{
AgoData * data = NULL;
if (agoIsValidContext(context)) {
if (memory_type == VX_MEMORY_TYPE_HOST) {
char desc[128]; sprintf(desc, "image:%4.4s,%d,%d", FORMAT_STR(color), addrs[0].dim_x, addrs[0].dim_y);
data = agoCreateDataFromDescription(context, NULL, desc, true);
if (data) {
agoGenerateDataName(context, "image-host", data->name);
agoAddData(&context->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&context->dataList, data->children[i]);
}
}
// set host allocated pointers
if (data->children) {
data->import_type = VX_MEMORY_TYPE_HOST;
for (vx_uint32 i = 0; i < data->numChildren; i++) {
data->children[i]->import_type = VX_MEMORY_TYPE_HOST;
data->children[i]->buffer = (vx_uint8 *)(ptrs ? ptrs[i] : nullptr);
data->children[i]->u.img.stride_in_bytes = addrs[i].stride_y;
data->children[i]->gpu_buffer_offset = 0;
#if (ENABLE_OPENCL || ENABLE_HIP)
data->children[i]->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
data->children[i]->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
#endif
}
}
else {
data->import_type = VX_MEMORY_TYPE_HOST;
data->buffer = (vx_uint8 *)(ptrs ? ptrs[0] : nullptr);
data->u.img.stride_in_bytes = addrs[0].stride_y;
data->gpu_buffer_offset = 0;
#if (ENABLE_OPENCL || ENABLE_HIP)
data->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
data->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
#endif
}
data->u.img.mem_handle = vx_false_e;
}
}
#if ENABLE_OPENCL
else if (memory_type == VX_MEMORY_TYPE_OPENCL) {
char desc[128]; sprintf(desc, "image:%4.4s,%d,%d", FORMAT_STR(color), addrs[0].dim_x, addrs[0].dim_y);
data = agoCreateDataFromDescription(context, NULL, desc, true);
if (data) {
agoGenerateDataName(context, "image-opencl", data->name);
agoAddData(&context->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&context->dataList, data->children[i]);
}
}
// set host allocated pointers
if (data->children) {
data->import_type = VX_MEMORY_TYPE_OPENCL;
for (vx_uint32 i = 0; i < data->numChildren; i++) {
data->children[i]->import_type = VX_MEMORY_TYPE_OPENCL;
data->children[i]->opencl_buffer = (cl_mem)(ptrs ? ptrs[i] : nullptr);
data->children[i]->gpu_buffer_offset = 0;
data->children[i]->u.img.stride_in_bytes = addrs[i].stride_y;
data->children[i]->gpu_buffer_offset = 0;
}
}
else {
data->import_type = VX_MEMORY_TYPE_OPENCL;
data->opencl_buffer = (cl_mem)(ptrs ? ptrs[0] : nullptr);
data->u.img.stride_in_bytes = addrs[0].stride_y;
data->gpu_buffer_offset = 0;
}
data->u.img.mem_handle = vx_false_e;
}
}
#elif ENABLE_HIP
else if (memory_type == VX_MEMORY_TYPE_HIP) {
char desc[128]; sprintf(desc, "image:%4.4s,%d,%d", FORMAT_STR(color), addrs[0].dim_x, addrs[0].dim_y);
data = agoCreateDataFromDescription(context, NULL, desc, true);
if (data) {
agoGenerateDataName(context, "image-hip", data->name);
agoAddData(&context->dataList, data);
// if data has children, add them too
if (data->children) {
for (vx_uint32 i = 0; i < data->numChildren; i++) {
agoAddData(&context->dataList, data->children[i]);
}
}
// set host allocated pointers
if (data->children) {
data->import_type = VX_MEMORY_TYPE_HIP;
for (vx_uint32 i = 0; i < data->numChildren; i++) {
data->children[i]->import_type = VX_MEMORY_TYPE_HIP;
data->children[i]->hip_memory = (vx_uint8 *)(ptrs ? ptrs[i] : nullptr);
data->children[i]->gpu_buffer_offset = 0;
data->children[i]->u.img.stride_in_bytes = addrs[i].stride_y;
data->children[i]->gpu_buffer_offset = 0;
}
}
else {
data->import_type = VX_MEMORY_TYPE_HIP;
data->hip_memory = (vx_uint8 *)(ptrs ? ptrs[0] : nullptr);
data->u.img.stride_in_bytes = addrs[0].stride_y;
data->gpu_buffer_offset = 0;
}
}
}
#endif
}
return (vx_image)data;
}
/*! \brief Swaps the image handle of an image previously created from handle.
*
* This function sets the new image handle (i.e. pointer to all image planes)
* and returns the previous one.
*
* Once this function call has completed, the application gets back the
* ownership of the memory referenced by the previous handle. This memory
* contains up-to-date pixel data, and the application can safely reuse or
* release it.
*
* The memory referenced by the new handle must have been allocated
* consistently with the image properties since the import type,
* memory layout and dimensions are unchanged (see addrs, color, and
* memory_type in \ref vxCreateImageFromHandle).
*
* All images created from ROI with this image as parent or ancestor
* will automatically use the memory referenced by the new handle.
*
* The behavior of SwapImageHandle when called from a user node is undefined.
* \param [in] image The reference to an image created from handle
* \param [in] new_ptrs[] pointer to a caller owned array that contains
* the new image handle (image plane pointers)
* \arg new_ptrs is non NULL. new_ptrs[i] must be non NULL for each i such as
* 0 < i < nbPlanes, otherwise, this is an error. The address of the storage memory
* for image plane i is set to new_ptrs[i]
* \arg new_ptrs is NULL: the previous image storage memory is reclaimed by the
* caller, while no new handle is provided.
* \param [out] prev_ptrs[] pointer to a caller owned array in which
* the application returns the previous image handle
* \arg prev_ptrs is non NULL. prev_ptrs must have at least as many
* elements as the number of image planes. For each i such as
* 0 < i < nbPlanes , prev_ptrs[i] is set to the address of the previous storage
* memory for plane i.
* \arg prev_ptrs NULL : the previous handle is not returned.
* \param [in] num_planes Number of planes in the image. This must be set equal to the number of planes of the input image.
* The number of elements in new_ptrs and prev_ptrs arrays must be equal to or greater than num_planes.
* If either array has more than num_planes elements, the extra elements are ignored. If either array is smaller
* than num_planes, the results are undefined.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_INVALID_REFERENCE image is not a valid image
* reference.
* \retval VX_ERROR_INVALID_PARAMETERS The image was not created from handle or
* the content of new_ptrs is not valid.
* \retval VX_FAILURE The image was already being accessed.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxSwapImageHandle(vx_image image_, void* const new_ptrs[], void* prev_ptrs[], vx_size num_planes)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
CAgoLock lock(image->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if(image->u.img.roiMasterImage) {
return status;
}
if (image->import_type == VX_MEMORY_TYPE_HOST && num_planes == image->u.img.planes) {
status = VX_SUCCESS;
if (image->children) {
for (vx_uint32 i = 0; i < image->numChildren; i++) {
if (prev_ptrs) prev_ptrs[i] = image->children[i]->buffer;
image->children[i]->buffer = (vx_uint8 *)(new_ptrs ? new_ptrs[i] : nullptr);
if (image->children[i]->buffer) {
image->children[i]->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->children[i]->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
}
// propagate to ROIs
for (auto roi = image->children[i]->roiDepList.begin(); roi != image->children[i]->roiDepList.end(); roi++) {
if(image->buffer) {
(*roi)->buffer = image->children[i]->buffer +
(*roi)->children[i]->u.img.rect_roi.start_y * (*roi)->children[i]->u.img.stride_in_bytes +
ImageWidthInBytesFloor((*roi)->children[i]->u.img.rect_roi.start_x, (*roi)->children[i]);
}
else {
(*roi)->buffer = nullptr;
}
}
image->children[i]->u.img.mem_handle = (new_ptrs == NULL) ? vx_true_e : vx_false_e;
}
}
else {
if (prev_ptrs) prev_ptrs[0] = image->buffer;
image->buffer = (vx_uint8 *)(new_ptrs ? new_ptrs[0] : nullptr);
if (image->buffer) {
image->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
}
// propagate to ROIs
for (auto roi = image->roiDepList.begin(); roi != image->roiDepList.end(); roi++) {
if(image->buffer) {
(*roi)->buffer = image->buffer +
(*roi)->u.img.rect_roi.start_y * image->u.img.stride_in_bytes +
ImageWidthInBytesFloor((*roi)->u.img.rect_roi.start_x, image);
}
else {
(*roi)->buffer = nullptr;
}
}
image->u.img.mem_handle = (new_ptrs == NULL) ? vx_true_e : vx_false_e;
}
}
#if ENABLE_OPENCL
else if (image->import_type == VX_MEMORY_TYPE_OPENCL && num_planes == image->u.img.planes) {
status = VX_SUCCESS;
if (image->children) {
for (vx_uint32 i = 0; i < image->numChildren; i++) {
if (prev_ptrs) prev_ptrs[i] = image->children[i]->opencl_buffer;
image->children[i]->opencl_buffer = (cl_mem)(new_ptrs ? new_ptrs[i] : nullptr);
if (image->children[i]->opencl_buffer) {
image->children[i]->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->children[i]->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL;
}
// propagate to ROIs
for (auto roi = image->children[i]->roiDepList.begin(); roi != image->children[i]->roiDepList.end(); roi++) {
(*roi)->opencl_buffer = image->children[i]->opencl_buffer;
}
image->children[i]->u.img.mem_handle = (new_ptrs == NULL) ? vx_true_e : vx_false_e;
}
}
else {
if (prev_ptrs) prev_ptrs[0] = image->opencl_buffer;
image->opencl_buffer = (cl_mem)(new_ptrs ? new_ptrs[0] : nullptr);
if (image->opencl_buffer) {
image->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL;
}
// propagate to ROIs
for (auto roi = image->roiDepList.begin(); roi != image->roiDepList.end(); roi++) {
(*roi)->opencl_buffer = image->opencl_buffer;
}
image->u.img.mem_handle = (new_ptrs == NULL) ? vx_true_e : vx_false_e;
}
}
#elif ENABLE_HIP
else if (image->import_type == VX_MEMORY_TYPE_HIP && num_planes == image->u.img.planes) {
status = VX_SUCCESS;
if (image->children) {
for (vx_uint32 i = 0; i < image->numChildren; i++) {
if (prev_ptrs) prev_ptrs[i] = (void *)image->children[i]->hip_memory;
image->children[i]->hip_memory = (vx_uint8 *)(new_ptrs ? new_ptrs[i] : nullptr);
if (image->children[i]->hip_memory) {
image->children[i]->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->children[i]->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL;
}
// propagate to ROIs
for (auto roi = image->children[i]->roiDepList.begin(); roi != image->children[i]->roiDepList.end(); roi++) {
(*roi)->hip_memory = image->children[i]->hip_memory;
}
}
}
else {
if (prev_ptrs) prev_ptrs[0] = image->hip_memory;
image->hip_memory = (vx_uint8 *)(new_ptrs ? new_ptrs[0] : nullptr);
if (image->hip_memory) {
image->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL;
}
// propagate to ROIs
for (auto roi = image->roiDepList.begin(); roi != image->roiDepList.end(); roi++) {
(*roi)->hip_memory = image->hip_memory;
}
}
}
#endif
}
return status;
}
/*! \brief Retrieves various attributes of an image.
* \param [in] image The reference to the image to query.
* \param [in] attribute The attribute to query. Use a \ref vx_image_attribute_e.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the image is not a \ref vx_image.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \retval VX_ERROR_NOT_SUPPORTED If the attribute is not supported on this implementation.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryImage(vx_image image_, vx_enum attribute, void *ptr, vx_size size)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
CAgoLock lock(image->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_IMAGE_WIDTH:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = image->u.img.width;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_HEIGHT:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = image->u.img.height;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_FORMAT:
if (size == sizeof(vx_uint32)) {
*(vx_df_image *)ptr = image->u.img.format;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_PLANES:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = image->u.img.planes;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_SPACE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = image->u.img.color_space;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_RANGE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = image->u.img.channel_range;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_MEMORY_TYPE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = image->import_type;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_SIZE:
if (size == sizeof(vx_size)) {
status = VX_SUCCESS;
if (image->numChildren) {
size = 0;
for (vx_uint32 plane = 0; plane < image->u.img.planes; plane++) {
if (!image->children[plane]->size) {
if (image->children[plane]->isNotFullyConfigured || agoDataSanityCheckAndUpdate(image->children[plane])) {
status = VX_ERROR_INVALID_REFERENCE;
}
}
size += image->children[plane]->size;
}
if (status == VX_SUCCESS)
*(vx_size *)ptr = size;
}
else {
if (!image->size) {
if (image->isNotFullyConfigured || agoDataSanityCheckAndUpdate(image)) {
status = VX_ERROR_INVALID_REFERENCE;
}
}
if (status == VX_SUCCESS)
*(vx_size *)ptr = image->size;
}
}
break;
#if ENABLE_OPENCL
case VX_IMAGE_ATTRIBUTE_AMD_OPENCL_BUFFER:
if (size == sizeof(cl_mem)) {
if (image->opencl_buffer) {
*(cl_mem *)ptr = image->opencl_buffer;
}
else {
#if defined(CL_VERSION_2_0)
*(vx_uint8 **)ptr = image->opencl_svm_buffer;
#else
*(vx_uint8 **)ptr = NULL;
#endif
}
status = VX_SUCCESS;
}
break;
case VX_IMAGE_ATTRIBUTE_AMD_GPU_BUFFER_OFFSET:
if (size == sizeof(cl_uint)) {
*(cl_uint *)ptr = image->gpu_buffer_offset;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_ATTRIBUTE_AMD_GPU_BUFFER_STRIDE:
if (size == sizeof(cl_uint)) {
*(cl_uint *)ptr = image->u.img.stride_in_bytes;
status = VX_SUCCESS;
}
break;
#elif ENABLE_HIP
case VX_IMAGE_ATTRIBUTE_AMD_HIP_BUFFER:
if (image->hip_memory) {
*(vx_uint8 **)ptr = image->hip_memory;
}
else {
*(vx_uint8 **)ptr = NULL;
}
status = VX_SUCCESS;
break;
case VX_IMAGE_ATTRIBUTE_AMD_GPU_BUFFER_OFFSET:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = image->gpu_buffer_offset;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_ATTRIBUTE_AMD_GPU_BUFFER_STRIDE:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = image->u.img.stride_in_bytes;
status = VX_SUCCESS;
}
break;
#endif
#if (ENABLE_OPENCL || ENABLE_HIP)
case VX_IMAGE_ATTRIBUTE_AMD_ENABLE_USER_BUFFER_GPU:
if (size == sizeof(vx_bool)) {
*(vx_bool *)ptr = image->u.img.enableUserBufferGPU;
status = VX_SUCCESS;
}
break;
#endif
case VX_IMAGE_ATTRIBUTE_AMD_HOST_BUFFER:
if (size == sizeof(vx_uint8)) {
if (image->buffer) {
vx_uint8** temp = static_cast< vx_uint8 **>(ptr);
*temp = image->buffer;
status = VX_SUCCESS;
}
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Allows setting attributes on the image.
* \param [in] image The reference to the image on which to set the attribute.
* \param [in] attribute The attribute to set. Use a \ref vx_image_attribute_e enumeration.
* \param [in] out The pointer to the location from which to read the value.
* \param [in] size The size of the object pointed to by \a out.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the image is not a \ref vx_image.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetImageAttribute(vx_image image_, vx_enum attribute, const void *ptr, vx_size size)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
CAgoLock lock(image->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_IMAGE_ATTRIBUTE_SPACE:
if (size == sizeof(vx_enum)) {
image->u.img.color_space = *(vx_color_space_e *)ptr;
status = VX_SUCCESS;
}
break;
case VX_IMAGE_ATTRIBUTE_RANGE:
if (size == sizeof(vx_enum)) {
image->u.img.channel_range = *(vx_channel_range_e *)ptr;
status = VX_SUCCESS;
}
break;
#if ENABLE_OPENCL
case VX_IMAGE_ATTRIBUTE_AMD_OPENCL_BUFFER:
if (size == sizeof(cl_mem) && image->u.img.enableUserBufferGPU) {
image->opencl_buffer = *(cl_mem *)ptr;
if (image->opencl_buffer) {
image->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL;
}
status = VX_SUCCESS;
}
break;
case VX_IMAGE_ATTRIBUTE_AMD_GPU_BUFFER_OFFSET:
if (size == sizeof(cl_uint) && image->u.img.enableUserBufferGPU) {
image->gpu_buffer_offset = *(cl_uint *)ptr;
status = VX_SUCCESS;
}
break;
#elif ENABLE_HIP
case VX_IMAGE_ATTRIBUTE_AMD_HIP_BUFFER:
if (image->u.img.enableUserBufferGPU) {
image->hip_memory = *(vx_uint8 **)ptr;
if (image->hip_memory) {
image->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
image->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL;
}
status = VX_SUCCESS;
}
break;
case VX_IMAGE_ATTRIBUTE_AMD_GPU_BUFFER_OFFSET:
if (size == sizeof(vx_uint32) && image->u.img.enableUserBufferGPU) {
image->gpu_buffer_offset = *(vx_uint32 *)ptr;
status = VX_SUCCESS;
}
break;
#endif
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Initialize an image with the given pixel value.
* \param [in] image The reference to the image to initialize.
* \param [in] pixel_value The pointer to the constant pixel value to initialize all image pixels. See \ref vx_pixel_value_t.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the image is a uniform image, a virtual image, or not a \ref vx_image.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \note All pixels of the entire image are initialized to the indicated pixel value, independently from the valid region.
* The valid region of the image is unaffected by this function. The image remains mutable after the call to this function,
* so its pixels and mutable attributes may be changed by subsequent functions.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetImagePixelValues(vx_image image, const vx_pixel_value_t *pixel_value)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
AgoData * image_data = (AgoData *)image;
if (agoIsValidData(image_data, VX_TYPE_IMAGE))
{
CAgoLock lock(image_data->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
vx_uint32 x, y, p, width, height;
vx_size planes = 0;
vx_rectangle_t rect = {0,0,0,0};
vxGetValidRegionImage(image, &rect);
vx_df_image format = 0;
vxQueryImage(image, VX_IMAGE_FORMAT, &format, sizeof(format));
vxQueryImage(image, VX_IMAGE_PLANES, &planes, sizeof(planes));
for (p = 0; p < planes; p++)
{
vx_imagepatch_addressing_t addr;
void *base = NULL;
if (vxAccessImagePatch(image, &rect, p, &addr, &base, VX_WRITE_ONLY) == VX_SUCCESS)
{
status = VX_SUCCESS;
width = (format == VX_DF_IMAGE_U1) ? addr.dim_x - rect.start_x % 8 : addr.dim_x;
height = addr.dim_y;
for (y = 0; y < height; y+=addr.step_y)
{
for (x = 0; x < width; x+=addr.step_x)
{
if (format == VX_DF_IMAGE_U1)
{
vx_uint32 xShftd = x + rect.start_x % 8;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, xShftd, y, &addr);
vx_uint8 offset = xShftd % 8;
vx_uint8 mask = 1 << offset;
*ptr = (*ptr & ~mask) | ((pixel_value->U1 ? 1 : 0) << offset);
}
else if (format == VX_DF_IMAGE_U8)
{
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel_value->U8;
}
else if (format == VX_DF_IMAGE_U16)
{
vx_uint16 *ptr = (vx_uint16*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel_value->U16;
}
else if (format == VX_DF_IMAGE_U32)
{
vx_uint32 *ptr = (vx_uint32*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel_value->U32;
}
else if (format == VX_DF_IMAGE_S16)
{
vx_int16 *ptr = (vx_int16*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel_value->S16;
}
else if (format == VX_DF_IMAGE_S32)
{
vx_int32 *ptr = (vx_int32*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel_value->S32;
}
else if ((format == VX_DF_IMAGE_RGB) ||
(format == VX_DF_IMAGE_RGBX))
{
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
ptr[0] = pixel_value->RGBX[0];
ptr[1] = pixel_value->RGBX[1];
ptr[2] = pixel_value->RGBX[2];
if (format == VX_DF_IMAGE_RGBX)
ptr[3] = pixel_value->RGBX[3];
}
else if ((format == VX_DF_IMAGE_YUV4) ||
(format == VX_DF_IMAGE_IYUV))
{
vx_uint8 *pixel = (vx_uint8 *)&pixel_value->YUV;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel[p];
}
else if ((p == 0) &&
((format == VX_DF_IMAGE_NV12) ||
(format == VX_DF_IMAGE_NV21)))
{
vx_uint8 *pixel = (vx_uint8 *)&pixel_value->YUV;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
*ptr = pixel[0];
}
else if ((p == 1) && (format == VX_DF_IMAGE_NV12))
{
vx_uint8 *pixel = (vx_uint8 *)&pixel_value->YUV;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
ptr[0] = pixel[1];
ptr[1] = pixel[2];
}
else if ((p == 1) && (format == VX_DF_IMAGE_NV21))
{
vx_uint8 *pixel = (vx_uint8 *)&pixel_value->YUV;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
ptr[0] = pixel[2];
ptr[1] = pixel[1];
}
else if (format == VX_DF_IMAGE_UYVY)
{
vx_uint8 *pixel = (vx_uint8 *)&pixel_value->YUV;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
if (x % 2 == 0)
{
ptr[0] = pixel[1];
ptr[1] = pixel[0];
}
else
{
ptr[0] = pixel[2];
ptr[1] = pixel[0];
}
}
else if (format == VX_DF_IMAGE_YUYV)
{
vx_uint8 *pixel = (vx_uint8 *)&pixel_value->YUV;
vx_uint8 *ptr = (vx_uint8*)vxFormatImagePatchAddress2d(base, x, y, &addr);
if (x % 2 == 0)
{
ptr[0] = pixel[0];
ptr[1] = pixel[1];
}
else
{
ptr[0] = pixel[0];
ptr[1] = pixel[2];
}
}
}
}
if (vxCommitImagePatch(image, &rect, p, &addr, base) != VX_SUCCESS)
{
status = VX_FAILURE;
break;
}
}
else
{
status = VX_FAILURE;
break;
}
}
}
return status;
}
/*! \brief Releases a reference to an image object.
* The object may not be garbage collected until its total reference count is zero.
* \param [in] image The pointer to the image to release.
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseImage(vx_image *image)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (image && agoIsValidData((AgoData*)*image, VX_TYPE_IMAGE)) {
if (!agoReleaseData((AgoData*)*image, true)) {
*image = NULL;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief This computes the size needed to retrieve an image patch from an image.
* \param [in] image The reference to the image from which to extract the patch.
* \param [in] rect The coordinates. Must be 0 <= start < end <= dimension where
* dimension is width for x and height for y.
* \param [in] plane_index The plane index from which to get the data.
* \return vx_size
* \ingroup group_image
*/
VX_API_ENTRY vx_size VX_API_CALL vxComputeImagePatchSize(vx_image image_,
const vx_rectangle_t *rect,
vx_uint32 plane_index)
{
AgoData * image = (AgoData *)image_;
vx_size size = 0;
if (agoIsValidData(image, VX_TYPE_IMAGE) && !image->isVirtual && rect && (plane_index < image->u.img.planes)) {
AgoData * img = image;
if (image->children) {
img = image->children[plane_index];
}
size = (vx_size)((unsigned long) ImageWidthInBytesFloor(((rect->end_x - rect->start_x) >> img->u.img.x_scale_factor_is_2), img) *
((rect->end_y - rect->start_y) >> img->u.img.y_scale_factor_is_2));
}
return size;
}
/*! \brief Allows the User to extract a rectangular patch (subset) of an image from a single plane.
* \param [in] image The reference to the image from which to extract the patch.
* \param [in] rect The coordinates from which to get the patch. Must be 0 <= start < end.
* \param [in] plane_index The plane index from which to get the data.
* \param [out] addr The addressing information for the image patch to be written into the data structure.
* \param [out] ptr The pointer to a pointer of a location to store the data.
* \arg If the user passes in a NULL, an error occurs.
* \arg If the user passes in a pointer to a NULL, the function returns internal memory, map, or allocates a buffer and returns it.
* \arg If the user passes in a pointer to a non-NULL pointer, the function attempts to
* copy to the location provided by the user.
*
* (*ptr) must be given to \ref vxCommitImagePatch.
* \param [in] usage This declares the intended usage of the pointer using the \ref vx_accessor_e enumeration.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_OPTIMIZED_AWAY The reference is a virtual image and cannot be accessed or committed.
* \retval VX_ERROR_INVALID_PARAMETERS The \a start, \a end, \a plane_index, \a stride_x, or \a stride_y pointer is incorrect.
* \retval VX_ERROR_INVALID_REFERENCE The image reference is not actually an image reference.
* \note The user may ask for data outside the bounds of the valid region, but
* such data has an undefined value.
* \note Users must be cautious to prevent passing in \e uninitialized pointers or
* addresses of uninitialized pointers to this function.
* \pre \ref vxComputeImagePatchSize if users wish to allocate their own memory.
* \post \ref vxCommitImagePatch with same (*ptr) value.
* \ingroup group_image
* \include vx_imagepatch.c
*/
VX_API_ENTRY vx_status VX_API_CALL vxAccessImagePatch(vx_image image_,
const vx_rectangle_t *rect,
vx_uint32 plane_index,
vx_imagepatch_addressing_t *addr,
void **ptr,
vx_enum usage)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (image->isVirtual && !image->buffer) {
status = VX_ERROR_OPTIMIZED_AWAY;
}
else if ((plane_index < image->u.img.planes) && addr && ptr && rect &&
rect->start_x < rect->end_x && rect->start_y < rect->end_y &&
rect->end_x <= image->u.img.width && rect->end_y <= image->u.img.height &&
(!image->u.img.isUniform || usage == VX_READ_ONLY) && !image->isNotFullyConfigured)
{
AgoData * img = image;
if (image->children) {
img = image->children[plane_index];
}
if (!img->buffer) {
CAgoLock lock(img->ref.context->cs);
if (agoAllocData(img)) {
return VX_ERROR_NO_MEMORY;
}
}
if (!*ptr) {
addr->dim_x = (rect->end_x - rect->start_x);
addr->dim_y = (rect->end_y - rect->start_y);
addr->scale_x = VX_SCALE_UNITY >> img->u.img.x_scale_factor_is_2;
addr->scale_y = VX_SCALE_UNITY >> img->u.img.y_scale_factor_is_2;
addr->step_x = 1 << img->u.img.x_scale_factor_is_2;
addr->step_y = 1 << img->u.img.y_scale_factor_is_2;
addr->stride_x = ((img->u.img.pixel_size_in_bits_num & 7) || (img->u.img.pixel_size_in_bits_denom > 1)) ?
0 : (img->u.img.pixel_size_in_bits_num >> 3);
addr->stride_y = img->u.img.stride_in_bytes;
}
vx_uint8 * ptr_internal = img->buffer +
(rect->start_y >> img->u.img.y_scale_factor_is_2) * img->u.img.stride_in_bytes +
ImageWidthInBytesFloor((rect->start_x >> img->u.img.x_scale_factor_is_2), img);
vx_uint8 * ptr_returned = *ptr ? (vx_uint8 *)*ptr : ptr_internal;
// save the pointer and usage for use in vxCommitImagePatch
status = VX_SUCCESS;
for (auto i = img->mapped.begin(); i != img->mapped.end(); i++) {
if (i->ptr == ptr_returned) {
// can't support vxAccessImagePatch() more than once with same pointer
// the application needs to call vxCommitImagePatch() before calling vxAccessImagePatch()
status = VX_FAILURE;
}
}
if (status == VX_SUCCESS) {
MappedData item = { img->nextMapId++, ptr_returned, usage, (ptr_returned != ptr_internal) ? true : false };
img->mapped.push_back(item);
*ptr = ptr_returned;
if (usage == VX_READ_ONLY || usage == VX_READ_AND_WRITE) {
#if ENABLE_OPENCL
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
if (dataToSync->opencl_buffer && !(dataToSync->buffer_sync_flags & AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED)) {
// make sure dirty OpenCL buffers are synched before giving access for read
if (dataToSync->buffer_sync_flags & (AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL)) {
cl_int err = clEnqueueReadBuffer(dataToSync->ref.context->opencl_cmdq, dataToSync->opencl_buffer, CL_TRUE, dataToSync->gpu_buffer_offset, dataToSync->size, dataToSync->buffer, 0, NULL, NULL);
if (err) {
status = VX_FAILURE;
agoAddLogEntry(&image->ref, status, "ERROR: vxAccessImagePatch: clEnqueueReadBuffer() => %d\n", err);
return status;
}
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED;
}
}
#elif ENABLE_HIP
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
if (dataToSync->hip_memory && !(dataToSync->buffer_sync_flags & AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED)) {
// make sure dirty OpenCL buffers are synched before giving access for read
// copy from Device to Host
if (dataToSync->buffer_sync_flags & (AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL)) {
hipError_t err = hipMemcpyDtoH((void *)dataToSync->buffer, (dataToSync->hip_memory + dataToSync->gpu_buffer_offset), dataToSync->size);
if (err) {
status = VX_FAILURE;
agoAddLogEntry(&image->ref, status, "ERROR: vxAccessImagePatch: hipMemcpyDtoH() => %d\n", err);
return status;
}
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED;
}
}
#endif
if (item.used_external_ptr) {
// copy if read is requested with explicit external buffer
if (addr->stride_x == 0 || ((addr->stride_x << 3) == img->u.img.pixel_size_in_bits_num && img->u.img.pixel_size_in_bits_denom == 1))
HafCpu_ChannelCopy_U8_U8(ImageWidthInBytesFloor((rect->end_x - rect->start_x) >> img->u.img.x_scale_factor_is_2, img),
((rect->end_y - rect->start_y) >> img->u.img.y_scale_factor_is_2), ptr_returned, addr->stride_y, ptr_internal, img->u.img.stride_in_bytes);
else
HafCpu_BufferCopyDisperseInDst(((rect->end_x - rect->start_x) >> img->u.img.x_scale_factor_is_2), ((rect->end_y - rect->start_y) >> img->u.img.y_scale_factor_is_2),
(img->u.img.pixel_size_in_bits_num / img->u.img.pixel_size_in_bits_denom + 7) >> 3, ptr_returned, addr->stride_y, addr->stride_x, ptr_internal, img->u.img.stride_in_bytes);
}
}
}
}
}
return status;
}
/*! \brief This allows the User to commit a rectangular patch (subset) of an image from a single plane.
* \param [in] image The reference to the image from which to extract the patch.
* \param [in] rect The coordinates to which to set the patch. Must be 0 <= start <= end.
* This may be 0 or a rectangle of zero area in order to indicate that the commit
* must only decrement the reference count.
* \param [in] plane_index The plane index to which to set the data.
* \param [in] addr The addressing information for the image patch.
* \param [in] ptr The pointer of a location from which to read the data. If the
* user allocated the pointer they must free it. If the pointer
* was set by \ref vxAccessImagePatch, the user may not access the pointer after
* this call completes.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_OPTIMIZED_AWAY The reference is a virtual image and cannot be accessed or committed.
* \retval VX_ERROR_INVALID_PARAMETERS The \a start, \a end, \a plane_index, \a stride_x, or \a stride_y pointer is incorrect.
* \retval VX_ERROR_INVALID_REFERENCE The image reference is not actually an image reference.
* \ingroup group_image
* \include vx_imagepatch.c
* \note If the implementation gives the client a pointer from
* \ref vxAccessImagePatch then implementation-specific behavior may occur.
* If not, then a copy occurs from the users pointer to the internal data of the object.
* \note If the rectangle intersects bounds of the current valid region, the
* valid region grows to the union of the two rectangles as long as they occur
* within the bounds of the original image dimensions.
*/
VX_API_ENTRY vx_status VX_API_CALL vxCommitImagePatch(vx_image image_,
const vx_rectangle_t *rect,
vx_uint32 plane_index,
const vx_imagepatch_addressing_t *addr,
const void *ptr)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
// check for ZERO AREA and mark rect as NULL for ZERO AREA
if (rect && ((rect->start_x == rect->end_x) || (rect->start_y == rect->end_y)))
rect = NULL;
// check for valid arguments
status = VX_ERROR_INVALID_PARAMETERS;
if (image->isVirtual && !image->buffer) {
status = VX_ERROR_OPTIMIZED_AWAY;
}
else if ((plane_index < image->u.img.planes) && addr && ptr &&
(!rect || (rect->start_x < rect->end_x && rect->start_y < rect->end_y && rect->end_x <= image->u.img.width && rect->end_y <= image->u.img.height)))
{
status = VX_SUCCESS;
AgoData * img = image;
if (image->children) {
img = image->children[plane_index];
}
if (!img->buffer) {
status = VX_FAILURE;
}
else if (!img->mapped.empty()) {
vx_enum usage = VX_READ_ONLY;
bool used_external_ptr = false;
for (auto i = img->mapped.begin(); i != img->mapped.end(); i++) {
if (i->ptr == ptr) {
if (rect) {
usage = i->usage;
used_external_ptr = i->used_external_ptr;
}
img->mapped.erase(i);
break;
}
}
if (usage == VX_WRITE_ONLY || usage == VX_READ_AND_WRITE) {
if (used_external_ptr) {
// copy from external buffer
vx_uint8 * buffer = img->buffer + (rect->start_y >> img->u.img.y_scale_factor_is_2) * img->u.img.stride_in_bytes +
ImageWidthInBytesFloor((rect->start_x >> img->u.img.x_scale_factor_is_2), img);
if (addr->stride_x == 0 || ((addr->stride_x << 3) == img->u.img.pixel_size_in_bits_num && img->u.img.pixel_size_in_bits_denom == 1))
HafCpu_ChannelCopy_U8_U8(ImageWidthInBytesFloor(((rect->end_x - rect->start_x) >> img->u.img.x_scale_factor_is_2), img),
((rect->end_y - rect->start_y) >> img->u.img.y_scale_factor_is_2), buffer, img->u.img.stride_in_bytes, (vx_uint8 *)ptr, addr->stride_y);
else
HafCpu_BufferCopyDisperseInSrc(((rect->end_x - rect->start_x) >> img->u.img.x_scale_factor_is_2) * addr->stride_x, ((rect->end_y - rect->start_y) >> img->u.img.y_scale_factor_is_2),
(img->u.img.pixel_size_in_bits_num / img->u.img.pixel_size_in_bits_denom + 7) >> 3, buffer, img->u.img.stride_in_bytes, (vx_uint8 *)ptr, addr->stride_y, addr->stride_x);
}
// update sync flags
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
dataToSync->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
}
}
}
}
return status;
}
/*!
* \brief Accesses a specific indexed pixel in an image patch.
* \param [in] ptr The base pointer of the patch as returned from \ref vxAccessImagePatch.
* \param [in] index The 0 based index of the pixel count in the patch. Indexes increase horizontally by 1 then wrap around to the next row.
* \param [in] addr The pointer to the addressing mode information returned from \ref vxAccessImagePatch.
* \return void * Returns the pointer to the specified pixel.
* \pre \ref vxAccessImagePatch
* \include vx_imagepatch.c
* \ingroup group_image
*/
VX_API_ENTRY void * VX_API_CALL vxFormatImagePatchAddress1d(void *ptr, vx_uint32 index, const vx_imagepatch_addressing_t *addr)
{
vx_uint8 *new_ptr = NULL;
if (ptr && index < addr->dim_x*addr->dim_y)
{
vx_uint32 x = index % addr->dim_x;
vx_uint32 y = index / addr->dim_x;
vx_uint32 offset = vxComputePatchOffset(x, y, addr);
new_ptr = (vx_uint8 *)ptr;
new_ptr = &new_ptr[offset];
}
return new_ptr;
}
/*!
* \brief Accesses a specific pixel at a 2d coordinate in an image patch.
* \param [in] ptr The base pointer of the patch as returned from \ref vxAccessImagePatch.
* \param [in] x The x dimension within the patch.
* \param [in] y The y dimension within the patch.
* \param [in] addr The pointer to the addressing mode information returned from \ref vxAccessImagePatch.
* \return void * Returns the pointer to the specified pixel.
* \pre \ref vxAccessImagePatch
* \include vx_imagepatch.c
* \ingroup group_image
*/
VX_API_ENTRY void * VX_API_CALL vxFormatImagePatchAddress2d(void *ptr, vx_uint32 x, vx_uint32 y, const vx_imagepatch_addressing_t *addr)
{
vx_uint8 *new_ptr = NULL;
if (ptr && x < addr->dim_x && y < addr->dim_y)
{
vx_uint32 offset = vxComputePatchOffset(x, y, addr);
new_ptr = (vx_uint8 *)ptr;
new_ptr = &new_ptr[offset];
}
return new_ptr;
}
/*! \brief Retrieves the valid region of the image as a rectangle.
* \details After the image is allocated but has not been written to this
* returns the full rectangle of the image so that functions do not have to manage
* a case for uninitialized data. The image still retains an uninitialized
* value, but once the image is written to via any means such as \ref vxCommitImagePatch,
* the valid region is altered to contain the maximum bounds of the written
* area.
* \param [in] image The image from which to retrieve the valid region.
* \param [out] rect The destination rectangle.
* \return vx_status
* \retval VX_ERROR_INVALID_REFERENCE Invalid image.
* \retval VX_ERROR_INVALID_PARAMETERS Invalid rect.
* \retval VX_STATUS Valid image.
* \note This rectangle can be passed directly to \ref vxAccessImagePatch to get
* the full valid region of the image. Modifications from \ref vxCommitImagePatch
* grows the valid region.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxGetValidRegionImage(vx_image image_, vx_rectangle_t *rect)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE))
{
status = VX_ERROR_INVALID_PARAMETERS;
if (rect) {
*rect = image->u.img.rect_valid;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Allows the application to copy a rectangular patch from/into an image object plane.
* \param [in] image The reference to the image object that is the source or the
* destination of the copy.
* \param [in] image_rect The coordinates of the image patch. The patch must be within
* the bounds of the image. (start_x, start_y) gives the coordinates of the topleft
* pixel inside the patch, while (end_x, end_y) gives the coordinates of the bottomright
* element out of the patch. Must be 0 <= start < end <= number of pixels in the image dimension.
* \param [in] image_plane_index The plane index of the image object that is the source or the
* destination of the patch copy.
* \param [in] user_addr The address of a structure describing the layout of the
* user memory location pointed by user_ptr. In the structure, only dim_x, dim_y,
* stride_x and stride_y fields must be provided, other fields are ignored by the function.
* The layout of the user memory must follow a row major order:
* stride_x >= pixel size in bytes, and stride_y >= stride_x * dim_x.
* \param [in] user_ptr The address of the memory location where to store the requested data
* if the copy was requested in read mode, or from where to get the data to store into the image
* object if the copy was requested in write mode. The accessible memory must be large enough
* to contain the specified patch with the specified layout:
* accessible memory in bytes >= (end_y - start_y) * stride_y.
* \param [in] usage This declares the effect of the copy with regard to the image object
* using the \ref vx_accessor_e enumeration. For uniform images, only VX_READ_ONLY
* is supported. For other images, Only VX_READ_ONLY and VX_WRITE_ONLY are supported:
* \arg VX_READ_ONLY means that data is copied from the image object into the application memory
* \arg VX_WRITE_ONLY means that data is copied into the image object from the application memory
* \param [in] user_mem_type A \ref vx_memory_type_e enumeration that specifies
* the memory type of the memory referenced by the user_addr.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_OPTIMIZED_AWAY This is a reference to a virtual image that cannot be
* accessed by the application.
* \retval VX_ERROR_INVALID_REFERENCE The image reference is not actually an image reference.
* \retval VX_ERROR_INVALID_PARAMETERS An other parameter is incorrect.
* \note The application may ask for data outside the bounds of the valid region, but
* such data has an undefined value.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxCopyImagePatch(vx_image image_, const vx_rectangle_t *image_rect, vx_uint32 image_plane_index, const vx_imagepatch_addressing_t *user_addr, void * user_ptr, vx_enum usage, vx_enum user_mem_type)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE))
{
status = VX_ERROR_INVALID_PARAMETERS;
if ((user_mem_type == VX_MEMORY_TYPE_HOST) && user_ptr && (usage == VX_READ_ONLY || usage == VX_WRITE_ONLY)) {
vx_rectangle_t rect = *image_rect;
vx_imagepatch_addressing_t addr = *user_addr;
status = vxAccessImagePatch(image_, &rect, image_plane_index, &addr, &user_ptr, usage);
if (status == VX_SUCCESS) {
status = vxCommitImagePatch(image_, &rect, image_plane_index, &addr, user_ptr);
}
}
}
return status;
}
/*! \brief Allows the application to get direct access to a rectangular patch of an image object plane.
* \param [in] image The reference to the image object that contains the patch to map.
* \param [in] rect The coordinates of image patch. The patch must be within the
* bounds of the image. (start_x, start_y) gives the coordinate of the topleft
* element inside the patch, while (end_x, end_y) give the coordinate of
* the bottomright element out of the patch. Must be 0 <= start < end.
* \param [in] plane_index The plane index of the image object to be accessed.
* \param [out] map_id The address of a vx_map_id variable where the function
* returns a map identifier.
* \arg (*map_id) must eventually be provided as the map_id parameter of a call to
* \ref vxUnmapImagePatch.
* \param [out] addr The address of a structure describing the memory layout of the
* image patch to access. The function fills the structure pointed by addr with the
* layout information that the application must consult to access the pixel data
* at address (*ptr). The layout of the mapped memory follows a row-major order:
* stride_x>0, stride_y>0 and stride_y >= stride_x * dim_x.
* If the image object being accessed was created via
* \ref vxCreateImageFromHandle, then the returned memory layout will be
* the identical to that of the addressing structure provided when
* \ref vxCreateImageFromHandle was called.
* \param [out] ptr The address of a pointer that the function sets to the
* address where the requested data can be accessed. This returned (*ptr) address
* is only valid between the call to this function and the corresponding call to
* \ref vxUnmapImagePatch.
* If image was created via \ref vxCreateImageFromHandle then the returned
* address (*ptr) will be the address of the patch in the original pixel buffer
* provided when image was created.
* \param [in] usage This declares the access mode for the image patch, using
* the \ref vx_accessor_e enumeration. For uniform images, only VX_READ_ONLY
* is supported.
* \arg VX_READ_ONLY: after the function call, the content of the memory location
* pointed by (*ptr) contains the image patch data. Writing into this memory location
* is forbidden and its behavior is undefined.
* \arg VX_READ_AND_WRITE : after the function call, the content of the memory
* location pointed by (*ptr) contains the image patch data; writing into this memory
* is allowed only for the location of pixels only and will result in a modification
* of the written pixels in the image object once the patch is unmapped. Writing into
* a gap between pixels (when addr->stride_x > pixel size in bytes or addr->stride_y > addr->stride_x*addr->dim_x)
* is forbidden and its behavior is undefined.
* \arg VX_WRITE_ONLY: after the function call, the memory location pointed by (*ptr)
* contains undefined data; writing each pixel of the patch is required prior to
* unmapping. Pixels not written by the application before unmap will become
* undefined after unmap, even if they were well defined before map. Like for
* VX_READ_AND_WRITE, writing into a gap between pixels is forbidden and its behavior
* is undefined.
* \param [in] mem_type A \ref vx_memory_type_e enumeration that
* specifies the type of the memory where the image patch is requested to be mapped.
* \param [in] flags An integer that allows passing options to the map operation.
* Use the \ref vx_map_flag_e enumeration.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_OPTIMIZED_AWAY This is a reference to a virtual image that cannot be
* accessed by the application.
* \retval VX_ERROR_INVALID_REFERENCE The image reference is not actually an image
* reference.
* \retval VX_ERROR_INVALID_PARAMETERS An other parameter is incorrect.
* \note The user may ask for data outside the bounds of the valid region, but
* such data has an undefined value.
* \ingroup group_image
* \post \ref vxUnmapImagePatch with same (*map_id) value.
*/
VX_API_ENTRY vx_status VX_API_CALL vxMapImagePatch(vx_image image_, const vx_rectangle_t *rect, vx_uint32 plane_index, vx_map_id *map_id, vx_imagepatch_addressing_t *addr, void **ptr, vx_enum usage, vx_enum mem_type, vx_uint32 flags)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (image->isVirtual && !image->buffer) {
status = VX_ERROR_OPTIMIZED_AWAY;
}
else if ((mem_type == VX_MEMORY_TYPE_HOST) && (plane_index < image->u.img.planes) && addr && ptr && rect &&
rect->start_x < rect->end_x && rect->start_y < rect->end_y &&
rect->end_x <= image->u.img.width && rect->end_y <= image->u.img.height &&
(!image->u.img.isUniform || usage == VX_READ_ONLY) && !image->isNotFullyConfigured)
{
AgoData * img = image;
if (image->children) {
img = image->children[plane_index];
}
if (!img->buffer) {
CAgoLock lock(img->ref.context->cs);
if (agoAllocData(img) || img->u.img.mem_handle) {
return VX_ERROR_NO_MEMORY;
}
}
vx_uint8 * ptr_returned = img->buffer +
(rect->start_y >> img->u.img.y_scale_factor_is_2) * img->u.img.stride_in_bytes +
ImageWidthInBytesFloor((rect->start_x >> img->u.img.x_scale_factor_is_2), img);
// save the pointer and usage for use in vxCommitImagePatch
status = VX_SUCCESS;
for (auto i = img->mapped.begin(); i != img->mapped.end(); i++) {
if (i->ptr == ptr_returned) {
// can't support vxAccessImagePatch() more than once with same pointer
// the application needs to call vxCommitImagePatch() before calling vxAccessImagePatch()
status = VX_FAILURE;
}
}
if (status == VX_SUCCESS) {
#if ENABLE_OPENCL
if (usage == VX_READ_ONLY || usage == VX_READ_AND_WRITE) {
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
if (dataToSync->opencl_buffer && !(dataToSync->buffer_sync_flags & AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED)) {
// make sure dirty OpenCL buffers are synched before giving access for read
if (dataToSync->buffer_sync_flags & (AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL)) {
cl_int err = clEnqueueReadBuffer(dataToSync->ref.context->opencl_cmdq, dataToSync->opencl_buffer, CL_TRUE, dataToSync->gpu_buffer_offset, dataToSync->size, dataToSync->buffer, 0, NULL, NULL);
if (err) {
status = VX_FAILURE;
agoAddLogEntry(&image->ref, status, "ERROR: vxMapImagePatch: clEnqueueReadBuffer() => %d\n", err);
return status;
}
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED;
}
}
}
else if (usage == VX_WRITE_ONLY)
{
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_WRITE;
}
#elif ENABLE_HIP
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
if (dataToSync->hip_memory && !(dataToSync->buffer_sync_flags & AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED)) {
// make sure dirty OpenCL buffers are synched before giving access for read
// copy from Device to Host
if (dataToSync->buffer_sync_flags & (AGO_BUFFER_SYNC_FLAG_DIRTY_BY_NODE_CL)) {
hipError_t err = hipMemcpyDtoH((void *)dataToSync->buffer, (dataToSync->hip_memory + dataToSync->gpu_buffer_offset), dataToSync->size);
if (err) {
status = VX_FAILURE;
agoAddLogEntry(&image->ref, status, "ERROR: vxAccessImagePatch: hipMemcpyDtoH() => %d\n", err);
return status;
}
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_SYNCHED;
}
}
if (usage == VX_WRITE_ONLY)
{
auto dataToSync = img->u.img.isROI ? img->u.img.roiMasterImage : img;
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_WRITE;
}
#endif
// get map id and set returned pointer
MappedData item = { img->nextMapId++, ptr_returned, usage, false, 0, plane_index };
image->mapped.push_back(item);
*map_id = item.map_id;
*ptr = ptr_returned;
addr->dim_x = rect->end_x - rect->start_x;
addr->dim_y = rect->end_y - rect->start_y;
addr->scale_x = VX_SCALE_UNITY >> img->u.img.x_scale_factor_is_2;
addr->scale_y = VX_SCALE_UNITY >> img->u.img.y_scale_factor_is_2;
addr->step_x = 1 << img->u.img.x_scale_factor_is_2;
addr->step_y = 1 << img->u.img.y_scale_factor_is_2;
addr->stride_x = (img->u.img.pixel_size_in_bits_denom > 1 || (img->u.img.pixel_size_in_bits_num & 7)) ? 0 : (img->u.img.pixel_size_in_bits_num >> 3);
addr->stride_y = img->u.img.stride_in_bytes;
}
}
}
return status;
}
/*! \brief Unmap and commit potential changes to a image object patch that were previously mapped.
* Unmapping an image patch invalidates the memory location from which the patch could
* be accessed by the application. Accessing this memory location after the unmap function
* completes has an undefined behavior.
* \param [in] image The reference to the image object to unmap.
* \param [out] map_id The unique map identifier that was returned by \ref vxMapImagePatch .
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_INVALID_REFERENCE The image reference is not actually an image reference.
* \retval VX_ERROR_INVALID_PARAMETERS An other parameter is incorrect.
* \ingroup group_image
* \pre \ref vxMapImagePatch with same map_id value
*/
VX_API_ENTRY vx_status VX_API_CALL vxUnmapImagePatch(vx_image image_, vx_map_id map_id)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE)) {
status = VX_ERROR_INVALID_PARAMETERS;
for (auto i = image->mapped.begin(); i != image->mapped.end(); i++) {
if (i->map_id == map_id) {
vx_enum usage = i->usage;
vx_uint32 plane = i->plane;
image->mapped.erase(i);
if (usage == VX_WRITE_ONLY || usage == VX_READ_AND_WRITE) {
// update sync flags
auto dataToSync = image->u.img.isROI ? image->u.img.roiMasterImage : image;
dataToSync->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
dataToSync->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
if (dataToSync->numChildren > 0 && plane < dataToSync->numChildren && dataToSync->children[plane]) {
dataToSync->children[plane]->buffer_sync_flags &= ~AGO_BUFFER_SYNC_FLAG_DIRTY_MASK;
dataToSync->children[plane]->buffer_sync_flags |= AGO_BUFFER_SYNC_FLAG_DIRTY_BY_COMMIT;
}
}
status = VX_SUCCESS;
break;
}
}
}
return status;
}
/*! \brief Create a sub-image from a single plane channel of another image.
*
* The sub-image refers to the data in the original image. Updates to this image
* update the parent image and reversely.
*
* The function supports only channels that occupy an entire plane of a multi-planar
* images, as listed below. Other cases are not supported.
* VX_CHANNEL_Y from YUV4, IYUV, NV12, NV21
* VX_CHANNEL_U from YUV4, IYUV, NV12, NV21
* VX_CHANNEL_V from YUV4, IYUV, NV12, NV21
*
* \param [in] img The reference to the parent image.
* \param [in] channel The \ref vx_channel_e channel to use.
* \returns An image reference \ref vx_image to the sub-image. Any possible errors preventing a
* successful creation should be checked using \ref vxGetStatus.
* \ingroup group_image
*/
VX_API_ENTRY vx_image VX_API_CALL vxCreateImageFromChannel(vx_image img, vx_enum channel)
{
AgoData * image = (AgoData *)img, * subImage = nullptr;
if (agoIsValidData(image, VX_TYPE_IMAGE))
{
if (image->numChildren > 0) {
if (channel == VX_CHANNEL_Y &&
(image->u.img.format == VX_DF_IMAGE_YUV4 || image->u.img.format == VX_DF_IMAGE_IYUV ||
image->u.img.format == VX_DF_IMAGE_NV12 || image->u.img.format == VX_DF_IMAGE_NV21))
{
subImage = image->children[0];
}
else if (channel == VX_CHANNEL_U && (image->u.img.format == VX_DF_IMAGE_YUV4 || image->u.img.format == VX_DF_IMAGE_IYUV))
{
subImage = image->children[1];
}
else if (channel == VX_CHANNEL_V && (image->u.img.format == VX_DF_IMAGE_YUV4 || image->u.img.format == VX_DF_IMAGE_IYUV))
{
subImage = image->children[2];
}
else if ((channel == VX_CHANNEL_U || channel == VX_CHANNEL_V) && (image->u.img.format == VX_DF_IMAGE_NV12 || image->u.img.format == VX_DF_IMAGE_NV21))
{
subImage = image->children[1];
}
}
}
if (subImage) {
subImage->ref.external_count++;
subImage->ref.context->num_active_references++;
}
return (vx_image)subImage;
}
/*! \brief Sets the valid rectangle for an image according to a supplied rectangle.
* \note Setting or changing the valid region from within a user node by means other than the call-back, for
* example by calling \ref vxSetImageValidRectangle, might result in an incorrect valid region calculation
* by the framework.
* \param [in] image The reference to the image.
* \param [in] rect The value to be set to the image valid rectangle. A NULL indicates that the valid region is the entire image.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE The image is not a \ref vx_image.
* \retval VX_ERROR_INVALID_PARAMETERS The rect does not define a proper valid rectangle.
* \ingroup group_image
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetImageValidRectangle(vx_image image_, const vx_rectangle_t *rect)
{
AgoData * image = (AgoData *)image_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(image, VX_TYPE_IMAGE) && !image->isVirtual)
{
status = VX_ERROR_INVALID_PARAMETERS;
if (!rect) {
image->u.img.rect_valid.start_x = 0;
image->u.img.rect_valid.start_y = 0;
image->u.img.rect_valid.end_x = image->u.img.width;
image->u.img.rect_valid.end_y = image->u.img.height;
status = VX_SUCCESS;
}
else if (rect->start_x < rect->end_x && rect->start_y < rect->end_y && rect->end_x <= image->u.img.width && rect->end_y <= image->u.img.height) {
image->u.img.rect_valid = *rect;
status = VX_SUCCESS;
}
if (status == VX_SUCCESS) {
// TBD: inform graphs for take this image as input, to re-computed output valid regions
}
}
return status;
}
/*==============================================================================
KERNEL
=============================================================================*/
/*! \brief Loads one or more kernels into the OpenVX context. This is the interface
* by which OpenVX is extensible. Once the set of kernels is loaded new kernels
* and their parameters can be queried.
* \note When all references to loaded kernels are released, the module
* may be automatically unloaded.
* \param [in] context The reference to the implementation context.
* \param [in] module The short name of the module to load. On systems where
* there are specific naming conventions for modules, the name passed
* should ignore such conventions. For example: \c libxyz.so should be
* passed as just \c xyz and the implementation will do the right thing that
* the platform requires.
* \note This API uses the system pre-defined paths for modules.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the context is not a \ref vx_context.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \ingroup group_user_kernels
* \see vxGetKernelByName
*/
VX_API_ENTRY vx_status VX_API_CALL vxLoadKernels(vx_context context, const vx_char *module)
{
return agoLoadModule(context, module);
}
/*! \brief Unloads all kernels from the OpenVX context that had been loaded from
* the module using the \ref vxLoadKernels function.
* \param [in] context The reference to the implementation context.
* \param [in] module The short name of the module to unload. On systems where
* there are specific naming conventions for modules, the name passed
* should ignore such conventions. For example: \c libxyz.so should be
* passed as just \c xyz and the implementation will do the right thing
* that the platform requires.
* \note This API uses the system pre-defined paths for modules.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the context is not a \ref
vx_context.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are
incorrect.
* \ingroup group_user_kernels
* \see vxLoadKernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxUnloadKernels(vx_context context, const vx_char *module)
{
return agoUnloadModule(context, module);
}
/*! \brief Registers a module with kernels in a context.
* \details This function registers the appropriate publish and unpublish functions
* with the module name if the module is not a dynamic library, so \ref vxLoadKernels and
* \ref vxUnloadKernels can be called.
* \param [in] context The reference to the context the kernels must be added to.
* \param [in] module The short name of the module to load.
* \param [in] publish must add kernels to the context by calling \ref vxAddUserKernel
* for each new kernel. It is called by \ref vxLoadKernels.
* \param [in] unpublish must remove kernels from the context by calling \ref vxRemoveKernel
* for each kernel the vxPublishKernels has added. It is called by \ref vxUnloadKernels.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors; any other value indicates failure.
* \retval VX_ERROR_INVALID_REFERENCE context is not a valid \ref vx_context reference.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \ingroup group_user_kernels
* \see vxLoadKernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxRegisterKernelLibrary(vx_context context, const vx_char *module, vx_publish_kernels_f publish, vx_unpublish_kernels_f unpublish)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidContext(context))
{
// TBD: add support to load obj files or static libs
status = VX_SUCCESS;
}
return status;
}
/*! \brief Obtains a reference to a kernel using a string to specify the name.
* \param [in] context The reference to the implementation context.
* \param [in] name The string of the name of the kernel to get.
* \return A kernel reference or zero if an error occurred.
* \retval 0 The kernel name is not found in the context.
* \ingroup group_kernel
* \pre \ref vxLoadKernels if the kernel is not provided by the
* OpenVX implementation.
* \note User Kernels should follow a "dotted" heirarchical syntax. For example:
* "com.company.example.xyz".
*/
VX_API_ENTRY vx_kernel VX_API_CALL vxGetKernelByName(vx_context context, const vx_char *name)
{
vx_kernel akernel = NULL;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
akernel = agoFindKernelByName(context, name);
if (akernel) {
akernel->ref.external_count++;
}
}
return akernel;
}
/*! \brief Obtains a reference to the kernel using the \ref vx_kernel_e enumeration.
* \details Enum values above the standard set are assumed to apply to
* loaded libraries.
* \param [in] context The reference to the implementation context.
* \param [in] kernel A value from \ref vx_kernel_e or a vendor or client-defined value.
* \return A \ref vx_kernel.
* \retval 0 The kernel enumeration is not found in the context.
* \ingroup group_kernel
* \pre \ref vxLoadKernels if the kernel is not provided by the
* OpenVX implementation.
*/
VX_API_ENTRY vx_kernel VX_API_CALL vxGetKernelByEnum(vx_context context, vx_enum kernel)
{
vx_kernel akernel = NULL;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
akernel = agoFindKernelByEnum(context, kernel);
if (akernel) {
akernel->ref.external_count++;
}
}
return akernel;
}
/*! \brief This allows the client to query the kernel to get information about
* the number of parameters, enum values, etc.
* \param [in] kernel The kernel reference to query.
* \param [in] attribute The attribute to query. Use a \ref vx_kernel_attribute_e.
* \param [out] ptr The pointer to the location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the kernel is not a \ref vx_kernel.
* \retval VX_ERROR_INVALID_PARAMETERS If any of the other parameters are incorrect.
* \retval VX_ERROR_NOT_SUPPORTED If the attribute value is not supported in this implementation.
* \ingroup group_kernel
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryKernel(vx_kernel kernel, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidKernel(kernel)) {
CAgoLock lock(kernel->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_KERNEL_ATTRIBUTE_PARAMETERS:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = kernel->argCount;
status = VX_SUCCESS;
}
break;
case VX_KERNEL_ATTRIBUTE_NAME:
if (ptr != NULL && size >= VX_MAX_KERNEL_NAME) {
strncpy((char *)ptr, kernel->name, size);
status = VX_SUCCESS;
}
break;
case VX_KERNEL_ATTRIBUTE_ENUM:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = kernel->id;
status = VX_SUCCESS;
}
break;
case VX_KERNEL_ATTRIBUTE_LOCAL_DATA_SIZE:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = kernel->localDataSize;
status = VX_SUCCESS;
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Release the reference to the kernel.
* The object may not be garbage collected until its total reference count is zero.
* \param [in] kernel The pointer to the kernel reference to release.
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
* \ingroup group_kernel
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseKernel(vx_kernel *kernel)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (kernel && agoIsValidKernel(*kernel)) {
if (!agoReleaseKernel(*kernel, true)) {
*kernel = NULL;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Allows users to add custom kernels to the known kernel
* database in OpenVX at run-time. This would primarily be used by the module function
* \c vxPublishKernels.
* \param [in] context The reference to the implementation context.
* \param [in] name The string to use to match the kernel.
* \param [in] enumeration The enumerated value of the kernel to be used by clients.
* \param [in] func_ptr The process-local function pointer to be invoked.
* \param [in] numParams The number of parameters for this kernel.
* \param [in] input The pointer to \ref vx_kernel_input_validate_f, which validates the
* input parameters to this kernel.
* \param [in] output The pointer to \ref vx_kernel_output_validate_f , which validates the
* output parameters to this kernel.
* \param [in] init The kernel initialization function.
* \param [in] deinit The kernel de-initialization function.
* \ingroup group_user_kernels
* \return \ref vx_kernel
* \retval 0 Indicates that an error occurred when adding the kernel.
* \retval * Kernel added to OpenVX.
*/
VX_API_ENTRY vx_kernel VX_API_CALL vxAddKernel(vx_context context,
const vx_char name[VX_MAX_KERNEL_NAME],
vx_enum enumeration,
vx_kernel_f func_ptr,
vx_uint32 numParams,
vx_kernel_input_validate_f input,
vx_kernel_output_validate_f output,
vx_kernel_initialize_f init,
vx_kernel_deinitialize_f deinit)
{
vx_kernel kernel = NULL;
if (agoIsValidContext(context) && numParams > 0 && numParams <= AGO_MAX_PARAMS && func_ptr && input && output) {
CAgoLock lock(context->cs);
// make sure there are no kernels with the same name
if (!agoFindKernelByEnum(context, enumeration) && !agoFindKernelByName(context, name)) {
kernel = new AgoKernel;
// initialize references
agoResetReference(&kernel->ref, VX_TYPE_KERNEL, context, NULL);
for (vx_uint32 index = 0; index < AGO_MAX_PARAMS; index++) {
agoResetReference(&kernel->parameters[index].ref, VX_TYPE_PARAMETER, kernel->ref.context, &kernel->ref);
kernel->parameters[index].scope = &kernel->ref;
}
// add kernel object to context
kernel->external_kernel = true;
kernel->ref.external_count++;
kernel->id = enumeration;
kernel->flags = AGO_KERNEL_FLAG_GROUP_USER | AGO_KERNEL_FLAG_DEVICE_CPU | AGO_KERNEL_FLAG_VALID_RECT_RESET;
strcpy(kernel->name, name);
kernel->argCount = numParams;
kernel->kernel_f = func_ptr;
kernel->input_validate_f = input;
kernel->output_validate_f = output;
kernel->initialize_f = init;
kernel->deinitialize_f = deinit;
kernel->importing_module_index_plus1 = context->importing_module_index_plus1;
kernel->user_kernel = vx_false_e;
agoAddKernel(&context->kernelList, kernel);
}
}
return kernel;
}
/*! \brief Allows users to add custom kernels to the known kernel
* database in OpenVX at run-time. This would primarily be used by the module function
* \ref vxPublishKernels.
* \param [in] context The reference to the implementation context.
* \param [in] name The string to use to match the kernel.
* \param [in] enumeration The enumerated value of the kernel to be used by clients.
* \param [in] func_ptr The process-local function pointer to be invoked.
* \param [in] numParams The number of parameters for this kernel.
* \param [in] validate The pointer to \ref vx_kernel_validate_f, which validates
* parameters to this kernel.
* \param [in] init The kernel initialization function.
* \param [in] deinit The kernel de-initialization function.
* \ingroup group_user_kernels
* \return \ref vx_kernel. Any possible errors
* preventing a successful creation should be checked using \ref vxGetStatus.
* \retval 0 Indicates that an error occurred when adding the kernel.
* \retval * Kernel added to OpenVX.
*/
VX_API_ENTRY vx_kernel VX_API_CALL vxAddUserKernel(vx_context context,
const vx_char name[VX_MAX_KERNEL_NAME],
vx_enum enumeration,
vx_kernel_f func_ptr,
vx_uint32 numParams,
vx_kernel_validate_f validate,
vx_kernel_initialize_f init,
vx_kernel_deinitialize_f deinit)
{
vx_kernel kernel = NULL;
if (agoIsValidContext(context) && numParams > 0 && numParams <= AGO_MAX_PARAMS && func_ptr && validate) {
CAgoLock lock(context->cs);
// make sure there are no kernels with the same name
if (!agoFindKernelByEnum(context, enumeration) && !agoFindKernelByName(context, name)) {
kernel = new AgoKernel;
// initialize references
agoResetReference(&kernel->ref, VX_TYPE_KERNEL, context, NULL);
for (vx_uint32 index = 0; index < AGO_MAX_PARAMS; index++) {
agoResetReference(&kernel->parameters[index].ref, VX_TYPE_PARAMETER, kernel->ref.context, &kernel->ref);
kernel->parameters[index].scope = &kernel->ref;
}
// add kernel object to context
kernel->external_kernel = true;
kernel->ref.external_count++;
kernel->id = enumeration;
kernel->flags = AGO_KERNEL_FLAG_GROUP_USER | AGO_KERNEL_FLAG_DEVICE_CPU | AGO_KERNEL_FLAG_VALID_RECT_RESET;
strcpy(kernel->name, name);
kernel->argCount = numParams;
kernel->kernel_f = func_ptr;
kernel->validate_f = validate;
kernel->initialize_f = init;
kernel->deinitialize_f = deinit;
kernel->importing_module_index_plus1 = context->importing_module_index_plus1;
kernel->user_kernel = vx_true_e;
agoAddKernel(&context->kernelList, kernel);
}
}
return kernel;
}
/*! \brief This API is called after all parameters have been added to the
* kernel and the kernel is \e ready to be used.
* \param [in] kernel The reference to the loaded kernel from \ref vxAddKernel.
* \return A \ref vx_status_e enumeration. If an error occurs, the kernel is not available
* for usage by the clients of OpenVX. Typically this is due to a mismatch
* between the number of parameters requested and given.
* \pre \ref vxAddKernel and \ref vxAddParameterToKernel
* \ingroup group_user_kernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxFinalizeKernel(vx_kernel kernel)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidKernel(kernel)) {
CAgoLock lock(kernel->ref.context->cs);
if (kernel->external_kernel && !kernel->finalized && kernel->argCount > 0) {
status = VX_SUCCESS;
// check if kernel has been initialized properly
for (vx_uint32 i = 0; i < kernel->argCount; i++) {
if (!kernel->argType[i] || !kernel->argConfig[i] || !kernel->parameters[i].scope) {
status = VX_ERROR_INVALID_REFERENCE;
break;
}
}
if (status == VX_SUCCESS) {
// mark that kernel has been finalized
kernel->finalized = true;
}
}
}
return status;
}
/*! \brief Allows users to set the signatures of the custom kernel.
* \param [in] kernel The reference to the kernel added with \ref vxAddKernel.
* \param [in] index The index of the parameter to add.
* \param [in] dir The direction of the parameter. This must be a value from \ref vx_direction_e.
* \param [in] data_type The type of parameter. This must be a value from \ref vx_type_e.
* \param [in] state The state of the parameter (required or not). This must be a value from \ref vx_parameter_state_e.
* \return A \ref vx_status_e enumerated value.
* \retval VX_SUCCESS Parameter is successfully set on kernel.
* \retval VX_ERROR_INVALID_REFERENCE The value passed as kernel was not a \c vx_kernel.
* \pre \ref vxAddKernel
* \ingroup group_user_kernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxAddParameterToKernel(vx_kernel kernel, vx_uint32 index, vx_enum dir, vx_enum data_type, vx_enum state)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidKernel(kernel)) {
CAgoLock lock(kernel->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
// add parameter if the kernel is not finalized and not a built-in kernel and not initialized earlier
if((data_type == VX_TYPE_DELAY && dir != VX_INPUT))
return VX_ERROR_INVALID_PARAMETERS;
if (kernel->external_kernel && !kernel->finalized &&
index < AGO_MAX_PARAMS &&
(dir == VX_INPUT || dir == VX_OUTPUT || dir == VX_BIDIRECTIONAL) &&
(state == VX_PARAMETER_STATE_REQUIRED || state == VX_PARAMETER_STATE_OPTIONAL))
{
status = VX_SUCCESS;
// save parameter details
kernel->parameters[index].index = index;
kernel->parameters[index].direction = (vx_direction_e)dir;
kernel->argConfig[index] = (dir == VX_INPUT) ? AGO_KERNEL_ARG_INPUT_FLAG :
((dir == VX_OUTPUT) ? AGO_KERNEL_ARG_OUTPUT_FLAG : (AGO_KERNEL_ARG_INPUT_FLAG | AGO_KERNEL_ARG_OUTPUT_FLAG));
kernel->parameters[index].type = data_type;
kernel->argType[index] = data_type;
kernel->parameters[index].state = (vx_parameter_state_e)state;
if (state == VX_PARAMETER_STATE_OPTIONAL)
kernel->argConfig[index] |= AGO_KERNEL_ARG_OPTIONAL_FLAG;
kernel->parameters[index].scope = &kernel->ref;
// update argument count
if (index >= kernel->argCount)
kernel->argCount = index + 1;
}
}
return status;
}
/*! \brief Removes a non-finalized \ref vx_kernel from the \ref vx_context.
* Once a \ref vx_kernel has been finalized it cannot be removed.
* \param [in] kernel The reference to the kernel to remove. Returned from \ref vxAddKernel.
* \note Any kernel enumerated in the base standard
* cannot be removed; only kernels added through \ref vxAddKernel can
* be removed.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_INVALID_REFERENCE If an invalid kernel is passed in.
* \retval VX_ERROR_INVALID_PARAMETER If a base kernel is passed in.
* \ingroup group_user_kernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxRemoveKernel(vx_kernel kernel)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidKernel(kernel)) {
status = VX_ERROR_INVALID_PARAMETERS;
// release if the kernel is not finalized and not a built-in kernel or user kernel with validate_f without external references
if (!kernel->finalized ||
(kernel->validate_f && kernel->external_kernel && (kernel->flags & AGO_KERNEL_FLAG_GROUP_USER) /*&&
kernel->ref.internal_count < 2 && kernel->ref.external_count == 0*/))
{
CAgoLock lock(kernel->ref.context->cs);
kernel->finalized = false;
if (!agoReleaseKernel(kernel, true)) {
status = VX_SUCCESS;
}
}
}
return status;
}
/*! \brief Sets kernel attributes.
* \param [in] kernel The reference to the kernel.
* \param [in] attribute The enumeration of the attributes. See \ref vx_kernel_attribute_e.
* \param [in] ptr The pointer to the location from which to read the attribute.
* \param [in] size The size of the data area indicated by \a ptr in bytes.
* \note After a kernel has been passed to \ref vxFinalizeKernel, no attributes
* can be altered.
* \return A \ref vx_status_e enumeration.
* \ingroup group_user_kernels
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetKernelAttribute(vx_kernel kernel, vx_enum attribute, const void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidKernel(kernel)) {
CAgoLock lock(kernel->ref.context->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_KERNEL_ATTRIBUTE_LOCAL_DATA_SIZE:
if (size == sizeof(vx_size)) {
kernel->localDataSize = *(vx_size *)ptr;
status = VX_SUCCESS;
}
break;
case VX_KERNEL_ATTRIBUTE_AMD_NODE_REGEN_CALLBACK:
if (size == sizeof(void *)) {
if (!kernel->finalized) {
*((void **)&kernel->regen_callback_f) = *(void **)ptr;
status = VX_SUCCESS;
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
#if (ENABLE_OPENCL || ENABLE_HIP)
case VX_KERNEL_ATTRIBUTE_AMD_QUERY_TARGET_SUPPORT:
if (size == sizeof(void *)) {
if (!kernel->finalized) {
*((void **)&kernel->query_target_support_f) = *(void **)ptr;
status = VX_SUCCESS;
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
case VX_KERNEL_ATTRIBUTE_AMD_OPENCL_CODEGEN_CALLBACK:
if (size == sizeof(void *)) {
if (!kernel->finalized) {
*((void **)&kernel->opencl_codegen_callback_f) = *(void **)ptr;
status = VX_SUCCESS;
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
case VX_KERNEL_ATTRIBUTE_AMD_OPENCL_GLOBAL_WORK_UPDATE_CALLBACK:
if (size == sizeof(void *)) {
if (!kernel->finalized) {
*((void **)&kernel->opencl_global_work_update_callback_f) = *(void **)ptr;
status = VX_SUCCESS;
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
case VX_KERNEL_ATTRIBUTE_AMD_GPU_BUFFER_ACCESS_ENABLE:
if (size == sizeof(vx_bool)) {
if (!kernel->finalized && !kernel->gpu_buffer_update_callback_f) {
kernel->opencl_buffer_access_enable = *(vx_bool *)ptr;
status = VX_SUCCESS;
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
case VX_KERNEL_ATTRIBUTE_AMD_GPU_BUFFER_UPDATE_CALLBACK:
if (size == sizeof(AgoKernelGpuBufferUpdateInfo)) {
if (!kernel->finalized) {
AgoKernelGpuBufferUpdateInfo * info = (AgoKernelGpuBufferUpdateInfo *)ptr;
if (info->gpu_buffer_update_param_index >= kernel->argCount ||
info->gpu_buffer_update_callback_f == nullptr ||
kernel->parameters[info->gpu_buffer_update_param_index].direction != VX_INPUT ||
kernel->parameters[info->gpu_buffer_update_param_index].type != VX_TYPE_IMAGE ||
kernel->parameters[info->gpu_buffer_update_param_index].state != VX_PARAMETER_STATE_REQUIRED)
{
// param index has to point to required input images only
status = VX_ERROR_INVALID_PARAMETERS;
}
else {
kernel->gpu_buffer_update_callback_f = info->gpu_buffer_update_callback_f;
kernel->gpu_buffer_update_param_index = info->gpu_buffer_update_param_index;
kernel->opencl_buffer_access_enable = vx_true_e;
status = VX_SUCCESS;
}
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
#endif
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Retrieves a \ref vx_parameter from a \ref vx_kernel.
* \param [in] kernel The reference to the kernel.
* \param [in] index The index of the parameter.
* \return A \ref vx_parameter.
* \retval 0 Either the kernel or index is invalid.
* \retval * The parameter reference.
* \ingroup group_parameter
*/
VX_API_ENTRY vx_parameter VX_API_CALL vxGetKernelParameterByIndex(vx_kernel kernel, vx_uint32 index)
{
vx_parameter parameter = NULL;
if (agoIsValidKernel(kernel) && index < kernel->argCount) {
parameter = &kernel->parameters[index];
parameter->ref.external_count++;
}
return parameter;
}
/*==============================================================================
GRAPH
=============================================================================*/
/*! \brief Creates an empty graph.
* \param [in] context The reference to the implementation context.
* \return A graph reference.
* \retval 0 if an error occurred.
* \ingroup group_graph
*/
VX_API_ENTRY vx_graph VX_API_CALL vxCreateGraph(vx_context context)
{
vx_graph graph = NULL;
if (agoIsValidContext(context)) {
graph = agoCreateGraph(context);
}
return graph;
}
/*! \brief Releases a reference to a graph.
* The object may not be garbage collected until its total reference count is zero.
* Once the reference count is zero, all node references in the graph are automatically
* released as well. Data referenced by those nodes may not be released as
* the user may have external references to the data.
* \param [in] graph The pointer to the graph to release.
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
* \ingroup group_graph
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseGraph(vx_graph *graph)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (graph && agoIsValidGraph(*graph)) {
if (!agoReleaseGraph(*graph)) {
*graph = NULL;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Verifies the state of the graph before it is executed.
* This is useful to catch programmer errors and contract errors. If not verified,
* the graph verifies before being processed.
* \pre Memory for data objects is not guarenteed to exist before
* this call. \post After this call data objects exist unless
* the implementation optimized them out.
* \param [in] graph The reference to the graph to verify.
* \return A status code for graphs with more than one error; it is
* undefined which error will be returned. Register a log callback using \ref vxRegisterLogCallback
* to receive each specific error in the graph.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
* \retval VX_ERROR_MULTIPLE_WRITERS If the graph contains more than one writer
* to any data object.
* \retval VX_ERROR_INVALID_NODE If a node in the graph is invalid or failed be created.
* \retval VX_ERROR_INVALID_GRAPH If the graph contains cycles or some other invalid topology.
* \retval VX_ERROR_INVALID_TYPE If any parameter on a node is given the wrong type.
* \retval VX_ERROR_INVALID_VALUE If any value of any parameter is out of bounds of specification.
* \retval VX_ERROR_INVALID_FORMAT If the image format is not compatible.
* \ingroup group_graph
* \see vxConvertReference
* \see vxProcessGraph
*/
VX_API_ENTRY vx_status VX_API_CALL vxVerifyGraph(vx_graph graph)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidGraph(graph)) {
CAgoLock lock(graph->cs);
CAgoLock lock2(graph->ref.context->cs);
// mark that graph is not verified and can't be executed
//graph->verified = vx_false_e;
graph->isReadyToExecute = vx_false_e;
graph->state = VX_GRAPH_STATE_UNVERIFIED;
// check to see if user requested for graph dump
vx_uint32 ago_graph_dump = 0;
char textBuffer[256];
if (agoGetEnvironmentVariable("AGO_DUMP_GRAPH", textBuffer, sizeof(textBuffer))) {
ago_graph_dump = atoi(textBuffer);
}
if (ago_graph_dump) {
agoWriteGraph(graph, NULL, 0, stdout, "*INPUT*");
}
// set the AGO_DEFAULT_TARGET if it is not set by the user
if (!agoGetEnvironmentVariable("AGO_DEFAULT_TARGET", textBuffer, sizeof(textBuffer))) {
#if ENABLE_OPENCL || ENABLE_HIP
agoSetEnvironmentVariable("AGO_DEFAULT_TARGET", "GPU");
#else
agoSetEnvironmentVariable("AGO_DEFAULT_TARGET", "CPU");
#endif
}
// verify graph per OpenVX specification
status = agoVerifyGraph(graph);
if (status == VX_SUCCESS) {
// run graph optimizer
if (agoOptimizeGraph(graph)) {
status = VX_FAILURE;
}
// initialize graph
else if (agoInitializeGraph(graph)) {
status = VX_FAILURE;
}
// graph is ready to execute
else {
graph->isReadyToExecute = vx_true_e;
}
graph->verified = vx_true_e;
graph->state = VX_GRAPH_STATE_VERIFIED;
}
if (ago_graph_dump) {
if (status == VX_SUCCESS) {
agoWriteGraph(graph, NULL, 0, stdout, "*FINAL*");
}
}
}
if (status == VX_SUCCESS)
graph->verified = vx_true_e;
else
graph->verified = vx_false_e;
return status;
}
/*! \brief This function causes the synchronous processing of a graph. If the graph
* has not been verified, then the implementation verifies the graph
* immediately. If verification fails this function returns a status
* identical to what \ref vxVerifyGraph would return. After
* the graph verfies successfully then processing occurs. If the graph was
* previously verified via \ref vxVerifyGraph or \ref vxProcessGraph
* then the graph is processed. This function blocks until the graph is completed.
* \param [in] graph The graph to execute.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Graph has been processed.
* \retval VX_FAILURE A catastrophic error occurred during processing.
* \retval * See \ref vxVerifyGraph.
* \pre \ref vxVerifyGraph must return \ref VX_SUCCESS before this function will pass.
* \ingroup group_graph
* \see vxVerifyGraph
*/
VX_API_ENTRY vx_status VX_API_CALL vxProcessGraph(vx_graph graph)
{
vx_status status = agoProcessGraph(graph);
return status;
}
/*! \brief Schedules a graph for future execution.
* \param [in] graph The graph to schedule.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_NO_RESOURCES The graph cannot be scheduled now.
* \retval VX_ERROR_NOT_SUFFICIENT The graph is not verified and has failed
forced verification.
* \retval VX_SUCCESS The graph has been scheduled.
* \pre \ref vxVerifyGraph must return \ref VX_SUCCESS before this function will pass.
* \ingroup group_graph
*/
VX_API_ENTRY vx_status VX_API_CALL vxScheduleGraph(vx_graph graph)
{
return agoScheduleGraph(graph);
}
/*! \brief Waits for a specific graph to complete.
* \param [in] graph The graph to wait on.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS The graph has completed.
* \retval VX_FAILURE The graph has not completed yet.
* \pre \ref vxScheduleGraph
* \ingroup group_graph
*/
VX_API_ENTRY vx_status VX_API_CALL vxWaitGraph(vx_graph graph)
{
return agoWaitGraph(graph);
}
/*! \brief Allows the user to query attributes of the Graph.
* \param [in] graph The reference to the created graph.
* \param [in] attribute The \ref vx_graph_attribute_e type needed.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \ingroup group_graph
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryGraph(vx_graph graph, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidGraph(graph)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
CAgoLock lock(graph->cs);
switch (attribute)
{
case VX_GRAPH_ATTRIBUTE_NUMNODES:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = graph->nodeList.count;
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_STATUS:
if (size == sizeof(vx_status)) {
*(vx_status *)ptr = graph->status;
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_PERFORMANCE:
if (size == sizeof(vx_perf_t)) {
agoPerfCopyNormalize(graph->ref.context, (vx_perf_t *)ptr, &graph->perf);
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_STATE:
if (size == sizeof(vx_enum)) {
*(vx_status *)ptr = graph->state;
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_NUMPARAMETERS:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = (vx_uint32)graph->parameters.size();
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_OPTIMIZER_FLAGS:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = graph->optimizer_flags;
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_AFFINITY:
if (size == sizeof(AgoTargetAffinityInfo_)) {
*(AgoTargetAffinityInfo_ *)ptr = graph->attr_affinity;
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_PERFORMANCE_INTERNAL_LAST:
if (size == sizeof(AgoGraphPerfInternalInfo)) {
#if ENABLE_OPENCL
// normalize all time units into nanoseconds
uint64_t num = 1000000000, denom = (uint64_t)agoGetClockFrequency();
((AgoGraphPerfInternalInfo *)ptr)->kernel_enqueue = graph->gpu_perf.kernel_enqueue * num / denom;
((AgoGraphPerfInternalInfo *)ptr)->kernel_wait = graph->gpu_perf.kernel_wait * num / denom;
((AgoGraphPerfInternalInfo *)ptr)->buffer_read = graph->gpu_perf.buffer_read * num / denom;
((AgoGraphPerfInternalInfo *)ptr)->buffer_write = graph->gpu_perf.buffer_write * num / denom;
#else
memset(ptr, 0, size);
#endif
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_PERFORMANCE_INTERNAL_AVG:
if (size == sizeof(AgoGraphPerfInternalInfo)) {
#if ENABLE_OPENCL
if (graph->perf.num > 0) {
// normalize all time units into nanoseconds
uint64_t num = 1000000000, denom = (uint64_t)agoGetClockFrequency();
((AgoGraphPerfInternalInfo *)ptr)->kernel_enqueue = (graph->gpu_perf_total.kernel_enqueue / graph->perf.num) * num / denom;
((AgoGraphPerfInternalInfo *)ptr)->kernel_wait = (graph->gpu_perf_total.kernel_wait / graph->perf.num) * num / denom;
((AgoGraphPerfInternalInfo *)ptr)->buffer_read = (graph->gpu_perf_total.buffer_read / graph->perf.num) * num / denom;
((AgoGraphPerfInternalInfo *)ptr)->buffer_write = (graph->gpu_perf_total.buffer_write / graph->perf.num) * num / denom;
}
else
#endif
{
memset(ptr, 0, size);
}
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_PERFORMANCE_INTERNAL_PROFILE:
status = agoGraphDumpPerformanceProfile(graph, (const char *)ptr);
break;
#if ENABLE_OPENCL
case VX_GRAPH_ATTRIBUTE_AMD_OPENCL_COMMAND_QUEUE:
if (size == sizeof(cl_command_queue)) {
*(cl_command_queue *)ptr = graph->opencl_cmdq;
status = VX_SUCCESS;
}
break;
#endif
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Allows the set to attributes on the Graph.
* \param [in] graph The reference to the graph.
* \param [in] attribute The \ref vx_graph_attribute_e type needed.
* \param [in] ptr The location from which to read the value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \ingroup group_graph
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetGraphAttribute(vx_graph graph, vx_enum attribute, const void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidGraph(graph)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
CAgoLock lock(graph->cs);
switch (attribute)
{
case VX_GRAPH_ATTRIBUTE_AMD_IMPORT_FROM_TEXT:
if (size == sizeof(AgoGraphImportInfo)) {
status = VX_SUCCESS;
AgoGraphImportInfo * info = (AgoGraphImportInfo *)ptr;
if (agoReadGraphFromString(graph, info->ref, info->num_ref, info->data_registry_callback_f, info->data_registry_callback_obj, info->text, info->dumpToConsole)) {
status = VX_FAILURE;
}
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_EXPORT_TO_TEXT:
if (size == sizeof(AgoGraphExportInfo)) {
status = VX_SUCCESS;
AgoGraphExportInfo * info = (AgoGraphExportInfo *)ptr;
FILE * fp = stdout;
if (strcmp(info->fileName, "stdout") != 0) {
fp = fopen(info->fileName, "w");
if (!fp) {
status = VX_FAILURE;
agoAddLogEntry(&graph->ref, status, "ERROR: vxSetGraphAttribute: unable to create: %s\n", info->fileName);
}
}
else if (agoWriteGraph(graph, info->ref, info->num_ref, fp, info->comment)) {
status = VX_FAILURE;
}
if (fp && fp != stdout) {
fclose(fp);
}
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_OPTIMIZER_FLAGS:
if (size == sizeof(vx_uint32)) {
graph->optimizer_flags = *(vx_uint32 *)ptr;
status = VX_SUCCESS;
}
break;
case VX_GRAPH_ATTRIBUTE_AMD_AFFINITY:
if (size == sizeof(AgoTargetAffinityInfo_)) {
status = VX_SUCCESS;
graph->attr_affinity = *(AgoTargetAffinityInfo_ *)ptr;
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Adds the given parameter extracted from a \ref vx_node to the graph.
* \param [in] graph The graph reference that contains the node.
* \param [in] parameter The parameter reference to add to the graph from the node.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Parameter added to Graph.
* \retval VX_ERROR_INVALID_REFERENCE The parameter is not a valid \ref vx_parameter.
* \retval VX_ERROR_INVALID_PARAMETER The parameter is of a node not in this
* graph.
* \ingroup group_graph_parameters
*/
VX_API_ENTRY vx_status VX_API_CALL vxAddParameterToGraph(vx_graph graph, vx_parameter parameter)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidGraph(graph) && !graph->verified) {
status = VX_ERROR_INVALID_PARAMETERS;
if (!parameter || (agoIsValidParameter(parameter) && parameter->scope->type == VX_TYPE_NODE)) {
graph->parameters.push_back(parameter);
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Sets a reference to the parameter on the graph. The implementation
* must set this parameter on the originating node as well.
* \param [in] graph The graph reference.
* \param [in] index The parameter index.
* \param [in] value The reference to set to the parameter.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Parameter set to Graph.
* \retval VX_ERROR_INVALID_REFERENCE The value is not a valid \ref vx_reference.
* \retval VX_ERROR_INVALID_PARAMETER The parameter index is out of bounds or the
* dir parameter is incorrect.
* \ingroup group_graph_parameters
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetGraphParameterByIndex(vx_graph graph, vx_uint32 index, vx_reference value)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidGraph(graph) && !graph->verified) {
status = VX_ERROR_INVALID_PARAMETERS;
if ((index < graph->parameters.size()) && graph->parameters[index] && (!value || agoIsValidReference(value))) {
vx_parameter parameter = graph->parameters[index];
if (((vx_node)parameter->scope)->paramList[parameter->index]) {
agoReleaseData(((vx_node)parameter->scope)->paramList[parameter->index], false);
}
((vx_node)parameter->scope)->paramList[parameter->index] = (AgoData *)value;
if (((vx_node)parameter->scope)->paramList[parameter->index]) {
agoRetainData(graph, ((vx_node)parameter->scope)->paramList[parameter->index], false);
}
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Retrieves a \ref vx_parameter from a \ref vx_graph.
* \param [in] graph The graph.
* \param [in] index The index of the parameter.
* \return \ref vx_parameter reference.
* \retval 0 if the index is out of bounds.
* \retval * The parameter reference.
* \ingroup group_graph_parameters
*/
VX_API_ENTRY vx_parameter VX_API_CALL vxGetGraphParameterByIndex(vx_graph graph, vx_uint32 index)
{
vx_parameter parameter = NULL;
if (agoIsValidGraph(graph) && (index < graph->parameters.size())) {
parameter = graph->parameters[index];
parameter->ref.external_count++;
}
return parameter;
}
/*! \brief Returns a Boolean to indicate the state of graph verification.
* \param [in] graph The reference to the graph to check.
* \return A \ref vx_bool value.
* \retval vx_true_e The graph is verified.
* \retval vx_false_e The graph is not verified. It must be verified before
* execution either through \ref vxVerifyGraph or automatically through
* \ref vxProcessGraph or \ref vxScheduleGraph.
* \ingroup group_graph
*/
VX_API_ENTRY vx_bool VX_API_CALL vxIsGraphVerified(vx_graph graph)
{
vx_bool verified = vx_false_e;
if (agoIsValidGraph(graph)) {
verified = graph->verified ? vx_true_e : vx_false_e;
}
return verified;
}
/*==============================================================================
NODE
=============================================================================*/
/*! \brief Creates a reference to a node object for a given kernel.
* \details This node has no references assigned as parameters after completion.
* The client is then required to set these parameters manually by \ref vxSetParameterByIndex.
* When clients supply their own node creation functions (for use with User Kernels), this is the API
* to use along with the parameter setting API.
* \param [in] graph The reference to the graph in which this node exists.
* \param [in] kernel The kernel reference to associate with this new node.
* \return vx_node
* \retval 0 The node failed to create.
* \retval * A node was created.
* \ingroup group_adv_node
* \post Call \ref vxSetParameterByIndex for as many parameters as needed to be set.
*/
VX_API_ENTRY vx_node VX_API_CALL vxCreateGenericNode(vx_graph graph, vx_kernel kernel)
{
vx_node node = NULL;
if (agoIsValidGraph(graph) && agoIsValidKernel(kernel) && /*!graph->verified &&*/ kernel->finalized) {
CAgoLock lock(graph->cs);
// graph->reverify = graph->verified;
// graph->verified = vx_false_e;
// graph->state = VX_GRAPH_STATE_UNVERIFIED;
node = agoCreateNode(graph, kernel);
node->ref.external_count++;
}
return node;
}
/*! \brief Allows a user to query information out of a node.
* \param [in] node The reference to the node to query.
* \param [in] attribute Use \ref vx_node_attribute_e value to query for information.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Successful
* \retval VX_ERROR_INVALID_PARAMETERS The type or size is incorrect.
* \ingroup group_node
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryNode(vx_node node, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidNode(node)) {
CAgoLock lock(((vx_graph)node->ref.scope)->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_NODE_STATUS:
if (size == sizeof(vx_status)) {
*(vx_status *)ptr = node->status;
status = VX_SUCCESS;
}
break;
case VX_NODE_PERFORMANCE:
if (size == sizeof(vx_perf_t)) {
vx_perf_t * perf = &node->perf;
if (node->perf.num == 0) {
// TBD: need mapping of node performance into its subsets or superset
// For now, nodes that doesn't exist in the graph will report the overall graph
// performance because the nodes might have got morphed into other nodes have
// no accountability
perf = &((AgoGraph *)node->ref.scope)->perf;
}
agoPerfCopyNormalize(node->ref.context, (vx_perf_t *)ptr, perf);
status = VX_SUCCESS;
}
break;
case VX_NODE_BORDER:
if (size == sizeof(vx_border_mode_t) || size == sizeof(vx_border_t)) {
*(vx_border_mode_t *)ptr = node->attr_border_mode;
status = VX_SUCCESS;
}
break;
case VX_NODE_LOCAL_DATA_SIZE:
if (size == sizeof(vx_size)) {
*(vx_size *)ptr = node->localDataSize;
status = VX_SUCCESS;
}
break;
case VX_NODE_LOCAL_DATA_PTR:
if (size == sizeof(void *)) {
*(void **)ptr = node->localDataPtr;
status = VX_SUCCESS;
}
break;
case VX_NODE_PARAMETERS:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = node->paramCount;
status = VX_SUCCESS;
}
break;
case VX_NODE_VALID_RECT_RESET:
if (size == sizeof(vx_bool)) {
*(vx_bool *)ptr = node->valid_rect_reset;
status = VX_SUCCESS;
}
break;
case VX_NODE_ATTRIBUTE_AMD_AFFINITY:
if (size == sizeof(AgoTargetAffinityInfo_)) {
*(AgoTargetAffinityInfo_ *)ptr = node->attr_affinity;
status = VX_SUCCESS;
}
break;
#if ENABLE_OPENCL
case VX_NODE_ATTRIBUTE_AMD_OPENCL_COMMAND_QUEUE:
if (size == sizeof(cl_command_queue)) {
AgoGraph * graph = (AgoGraph *)node->ref.scope;
*(cl_command_queue *)ptr = graph->opencl_cmdq;
status = VX_SUCCESS;
}
break;
#elif ENABLE_HIP
case VX_NODE_ATTRIBUTE_AMD_HIP_STREAM:
if (size == sizeof(hipStream_t)){
AgoGraph * graph = (AgoGraph *)node->ref.scope;
*(hipStream_t *)ptr = graph->hip_stream0;
status = VX_SUCCESS;
}
break;
#endif
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Allows a user to set attribute of a node before Graph Validation.
* \param [in] node The reference to the node to set.
* \param [in] attribute Use \ref vx_node_attribute_e value to query for information.
* \param [out] ptr The output pointer to where to send the value.
* \param [in] size The size of the objects to which \a ptr points.
* \note Some attributes are inherited from the \ref vx_kernel, which was used
* to create the node. Some of these can be overridden using this API, notably
* \ref VX_NODE_ATTRIBUTE_LOCAL_DATA_SIZE and \ref VX_NODE_ATTRIBUTE_LOCAL_DATA_PTR.
* \ingroup group_node
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS The attribute was set.
* \retval VX_ERROR_INVALID_REFERENCE node is not a vx_node.
* \retval VX_ERROR_INVALID_PARAMETER size is not correct for the type needed.
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetNodeAttribute(vx_node node, vx_enum attribute, const void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidNode(node)) {
CAgoLock lock(((vx_graph)node->ref.scope)->cs);
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_NODE_ATTRIBUTE_BORDER_MODE:
if (size == sizeof(vx_border_mode_t) || size == sizeof(vx_border_t)) {
node->attr_border_mode = *(vx_border_mode_t *)ptr;
status = VX_SUCCESS;
}
break;
case VX_NODE_ATTRIBUTE_LOCAL_DATA_SIZE:
if (node->local_data_change_is_enabled) {
if (size == sizeof(vx_size)) {
node->localDataSize = *(vx_size *)ptr;
node->local_data_set_by_implementation = vx_false_e;
status = VX_SUCCESS;
}
}
else
{
status = VX_ERROR_NOT_SUPPORTED;
}
break;
case VX_NODE_ATTRIBUTE_LOCAL_DATA_PTR:
if(node->local_data_change_is_enabled) {
if (size == sizeof(void *)) {
node->localDataPtr = *(vx_uint8 **)ptr;
node->local_data_set_by_implementation = vx_false_e;
status = VX_SUCCESS;
}
}
else
{
status = VX_ERROR_NOT_SUPPORTED;
}
break;
case VX_NODE_ATTRIBUTE_AMD_AFFINITY:
if (size == sizeof(AgoTargetAffinityInfo_)) {
node->attr_affinity = *(AgoTargetAffinityInfo_ *)ptr;
status = VX_SUCCESS;
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Releases a reference to a Node object.
* The object may not be garbage collected until its total reference count is zero.
* \param [in] node The pointer to the reference of the node to release.
* \ingroup group_node
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseNode(vx_node *node)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (node && agoIsValidNode(*node)) {
if (!agoReleaseNode(*node)) {
*node = NULL;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Removes a Node from its parent Graph and releases it.
* \param [in] node The pointer to the node to remove and release.
* \ingroup group_node
* \post After returning from this function the reference is zeroed.
*/
VX_API_ENTRY vx_status VX_API_CALL vxRemoveNode(vx_node *node)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (node && agoIsValidNode(*node)) {
vx_node anode = *node;
vx_graph graph = (vx_graph)anode->ref.scope;
CAgoLock lock(graph->cs);
if (!graph->verified && anode->ref.external_count == 1) {
// only remove the kernels that are created externally
if (agoRemoveNode(&graph->nodeList, anode, true)) {
status = VX_FAILURE;
agoAddLogEntry(&anode->ref, status, "ERROR: vxRemoveNode: failed for %s\n", anode->akernel->name);
}
else {
*node = NULL;
status = VX_SUCCESS;
}
}
}
return status;
}
/*! \brief Assigns a callback to a node.
* If a callback already exists in this node, this function must return an error
* and the user may clear the callback by passing a NULL pointer as the callback.
* \param [in] node The reference to the node.
* \param [in] callback The callback to associate with completion of this
* specific node.
* \warning This must be used with extreme caution as it can \e ruin
* optimizations in the power/performance efficiency of a graph.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Callback assigned.
* \retval VX_ERROR_INVALID_REFERENCE The value passed as node was not a \ref vx_node.
* \ingroup group_node_callback
*/
VX_API_ENTRY vx_status VX_API_CALL vxAssignNodeCallback(vx_node node, vx_nodecomplete_f callback)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidNode(node)) {
node->callback = callback;
if (node->newchildnode) {
node->newchildnode->callback = callback;
}
status = VX_SUCCESS;
}
return status;
}
/*! \brief Retrieves the current node callback function pointer set on the node.
* \param [in] node The reference to the \ref vx_node object.
* \ingroup group_node_callback
* \return vx_nodecomplete_f The pointer to the callback function.
* \retval NULL No callback is set.
* \retval * The node callback function.
*/
VX_API_ENTRY vx_nodecomplete_f VX_API_CALL vxRetrieveNodeCallback(vx_node node)
{
vx_nodecomplete_f callback = NULL;
if (agoIsValidNode(node)) {
callback = node->callback;
}
return callback;
}
/*! \brief Sets the node target to the provided value. A success invalidates the graph
* that the node belongs to (\ref vxVerifyGraph must be called before the next execution)
* \param [in] node The reference to the \ref vx_node object.
* \param [in] target_enum The target enum to be set to the \ref vx_node object.
* Use a \ref vx_target_e.
* \param [in] target_string The target name ASCII string. This contains a valid value
* when target_enum is set to \ref VX_TARGET_STRING, otherwise it is ignored.
* \ingroup group_node
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS Node target set.
* \retval VX_ERROR_INVALID_REFERENCE If node is not a \ref vx_node.
* \retval VX_ERROR_NOT_SUPPORTED If the node kernel is not supported by the specified target.
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetNodeTarget(vx_node node, vx_enum target_enum, const char* target_string)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidNode(node)) {
status = VX_ERROR_NOT_SUPPORTED;
if (target_enum == VX_TARGET_ANY) {
status = VX_SUCCESS;
}
else if (target_enum == VX_TARGET_STRING) {
if (!target_string) {
status = VX_ERROR_INVALID_REFERENCE;
}
else if (!_stricmp(target_string, "any")) {
status = VX_SUCCESS;
}
else if (!_stricmp(target_string, "cpu")) {
if (node->attr_affinity.device_type == 0) {
node->attr_affinity.device_type = AGO_TARGET_AFFINITY_CPU;
status = VX_SUCCESS;
}
}
else if (!_stricmp(target_string, "gpu")) {
if (node->attr_affinity.device_type == 0) {
node->attr_affinity.device_type = AGO_TARGET_AFFINITY_GPU;
status = VX_SUCCESS;
}
}
}
}
return status;
}
/*! \brief Creates replicas of the same node first_node to process a set of objects
* stored in \ref vx_pyramid or \ref vx_object_array.
* first_node needs to have as parameter levels 0 of a \ref vx_pyramid or the index 0 of a \ref vx_object_array.
* Replica nodes are not accessible by the application through any means. An application request for removal of
* first_node from the graph will result in removal of all replicas. Any change of parameter or attribute of
* first_node will be propagated to the replicas. \ref vxVerifyGraph shall enforce consistency of parameters and attributes
* in the replicas.
* \param [in] graph The reference to the graph.
* \param [in] first_node The reference to the node in the graph that will be replicated.
* \param [in] replicate an array of size equal to the number of node parameters, vx_true_e for the parameters
* that should be iterated over (should be a reference to a vx_pyramid or a vx_object_array),
* vx_false_e for the parameters that should be the same across replicated nodes and for optional
* parameters that are not used. Should be vx_true_e for all output and bidirectional parameters.
* \param [in] number_of_parameters number of elements in the replicate array
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If the first_node is not a \ref vx_node, or it is not the first child of a vx_pyramid.
* \retval VX_ERROR_NOT_COMPATIBLE At least one of replicated parameters is not of level 0 of a pyramid or at index 0 of an object array.
* \retval VX_FAILURE If the node does not belong to the graph, or the number of objects in the parent objects of inputs and output are not the same.
* \ingroup group_node
*/
VX_API_ENTRY vx_status VX_API_CALL vxReplicateNode(vx_graph graph, vx_node first_node, vx_bool replicate[], vx_uint32 number_of_parameters)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidGraph(graph) && agoIsValidNode(first_node)) {
status = VX_FAILURE;
if (first_node->ref.scope == &graph->ref && first_node->paramCount == number_of_parameters) {
status = VX_SUCCESS;
AgoData ** paramList = first_node->paramList;
vx_uint32 num_levels = 0;
for (vx_uint32 i = 0; i < number_of_parameters; i++) {
if (replicate[i]) {
if (paramList[i] && paramList[i]->parent && paramList[i]->siblingIndex == 0 && (paramList[i]->parent->ref.type == VX_TYPE_PYRAMID || paramList[i]->parent->ref.type == VX_TYPE_OBJECT_ARRAY)) {
if (num_levels != paramList[i]->parent->numChildren) {
if (num_levels == 0) {
num_levels = paramList[i]->parent->numChildren;
}
else {
status = VX_FAILURE;
break;
}
}
}
else {
status = VX_ERROR_NOT_COMPATIBLE;
break;
}
}
}
if (num_levels < 2)
status = VX_ERROR_NOT_COMPATIBLE;
for (vx_uint32 level = 1; level < num_levels && status == VX_SUCCESS; level++) {
vx_node node = vxCreateGenericNode(graph, first_node->akernel);
status = vxGetStatus((vx_reference)node);
if (status == VX_SUCCESS) {
for (vx_uint32 i = 0; i < number_of_parameters && status == VX_SUCCESS; i++) {
if (replicate[i]) {
AgoData * param = paramList[i]->parent->children[level];
if (param) {
status = vxSetParameterByIndex(node, i, ¶mList[i]->parent->children[level]->ref);
}
else {
status = VX_FAILURE;
}
}
else if (paramList[i]) {
status = vxSetParameterByIndex(node, i, ¶mList[i]->ref);
}
}
}
}
}
}
return status;
}
/*==============================================================================
PARAMETER
=============================================================================*/
/*! \brief Retrieves a \ref vx_parameter from a \ref vx_node.
* \param [in] node The node from which to extract the parameter.
* \param [in] index The index of the parameter to which to get a reference.
* \return \ref vx_parameter
* \ingroup group_parameter
*/
VX_API_ENTRY vx_parameter VX_API_CALL vxGetParameterByIndex(vx_node node, vx_uint32 index)
{
vx_parameter parameter = NULL;
if (agoIsValidNode(node) && (index < node->paramCount)) {
if (agoUpdateDelaySlots(node) == VX_SUCCESS) {
parameter = &node->parameters[index];
parameter->ref.external_count++;
}
}
return parameter;
}
/*! \brief Releases a reference to a parameter object.
* The object may not be garbage collected until its total reference count is zero.
* \param [in] param The pointer to the parameter to release.
* \ingroup group_parameter
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseParameter(vx_parameter *param)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (param && agoIsValidParameter(*param)) {
if ((*param)->ref.external_count > 0) {
(*param)->ref.external_count--;
*param = NULL;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Sets the specified parameter data for a kernel on the node.
* \param [in] node The node that contains the kernel.
* \param [in] index The index of the parameter desired.
* \param [in] value The reference to the parameter.
* \return A \ref vx_status_e enumeration.
* \ingroup group_parameter
* \see vxSetParameterByReference
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetParameterByIndex(vx_node node, vx_uint32 index, vx_reference value)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidNode(node)) {
status = VX_ERROR_INVALID_PARAMETERS;
vx_graph graph = (AgoGraph *)node->ref.scope;
if (graph->verified) {
status = VX_ERROR_NOT_SUPPORTED;
}
else if (node->parameters[index].state == VX_PARAMETER_STATE_REQUIRED && !value) {
status = VX_ERROR_INVALID_REFERENCE;
}
else if ((index < node->paramCount) && (!node->parameters[index].type || !value || node->parameters[index].type == value->type || node->parameters[index].type == VX_TYPE_REFERENCE)) {
if (node->paramList[index]) {
agoReleaseData(node->paramList[index], false);
}
AgoData * data = (AgoData *)value;
if (data && agoIsPartOfDelay(data)) {
// get the trace to delay object from original node parameter without vxAgeDelay changes
int siblingTrace[AGO_MAX_DEPTH_FROM_DELAY_OBJECT], siblingTraceCount = 0;
AgoData * delay = agoGetSiblingTraceToDelayForInit(data, siblingTrace, siblingTraceCount);
if (delay) {
// get the data
data = agoGetDataFromTrace(delay, siblingTrace, siblingTraceCount);
}
}
node->paramList[index] = node->paramListForAgeDelay[index] = data;
if (node->paramList[index]) {
agoRetainData((AgoGraph *)node->ref.scope, node->paramList[index], false);
}
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Associates a parameter reference and a data reference with a kernel
* on a node.
* \param [in] parameter The reference to the kernel parameter.
* \param [in] value The value to associate with the kernel parameter.
* \return A \ref vx_status_e enumeration.
* \ingroup group_parameter
* \see vxGetParameterByIndex
*/
VX_API_ENTRY vx_status VX_API_CALL vxSetParameterByReference(vx_parameter parameter, vx_reference value)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidParameter(parameter) && parameter->scope->type == VX_TYPE_NODE && parameter->ref.external_count > 0) {
vx_node node = (vx_node)parameter->scope;
vx_uint32 index = parameter->index;
status = vxSetParameterByIndex(node, index, value);
}
return status;
}
/*! \brief Allows the client to query a parameter to determine its meta-information.
* \param [in] param The reference to the parameter.
* \param [in] attribute The attribute to query. Use a \ref vx_parameter_attribute_e.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \ingroup group_parameter
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryParameter(vx_parameter param, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidParameter(param)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_PARAMETER_ATTRIBUTE_DIRECTION:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = param->direction;
status = VX_SUCCESS;
}
break;
case VX_PARAMETER_ATTRIBUTE_INDEX:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = param->index;
status = VX_SUCCESS;
}
break;
case VX_PARAMETER_ATTRIBUTE_TYPE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = param->type;
status = VX_SUCCESS;
}
break;
case VX_PARAMETER_ATTRIBUTE_STATE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = param->state;
status = VX_SUCCESS;
}
break;
case VX_PARAMETER_ATTRIBUTE_REF:
if (size == sizeof(vx_reference)) {
vx_node node = (vx_node)param->scope;
if (agoIsValidNode(node)) {
if (param->index < node->paramCount) {
vx_reference ref = (vx_reference)node->paramList[param->index];
*(vx_reference *)ptr = ref;
// TBD: handle optimized buffers and kernels
if (ref) {
ref->external_count++;
ref->context->num_active_references++;
}
status = VX_SUCCESS;
}
}
else {
status = VX_ERROR_NOT_SUPPORTED;
}
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*==============================================================================
SCALAR
=============================================================================*/
/*! \brief Creates a reference to a scalar object. Also see \ref sub_node_parameters.
* \param [in] context The reference to the system context.
* \param [in] data_type The \ref vx_type_e of the scalar. Must be greater than
* \ref VX_TYPE_INVALID and less than \ref VX_TYPE_SCALAR_MAX.
* \param [in] ptr The pointer to the initial value of the scalar.
* \ingroup group_scalar
* \return A \ref vx_scalar reference.
* \retval 0 The scalar could not be created.
* \retval * The scalar was created. Check for further errors with \ref vxGetStatus.
*/
VX_API_ENTRY vx_scalar VX_API_CALL vxCreateScalar(vx_context context, vx_enum data_type, const void *ptr)
{
AgoData * data = NULL;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
vx_size size = agoType2Size(context, data_type);
if(size > 0 || data_type == VX_TYPE_STRING_AMD) {
data = (AgoData *)vxCreateScalarWithSize(context, data_type, ptr, size);
}
}
return (vx_scalar)data;
}
/*! \brief Creates an opaque reference to a scalar object with no direct user access.
* \param [in] graph The reference to the parent graph.
* \param [in] data_type The type of data to hold. Must be greater than VX_TYPE_INVALID and less than or equal to VX_TYPE_VENDOR_STRUCT_END.
* Or must be a vx_enum returned from vxRegisterUserStruct.
* \return A scalar reference vx_scalar. Any possible errors preventing a successful creation should be checked using vxGetStatus.
*/
VX_API_ENTRY vx_scalar VX_API_CALL vxCreateVirtualScalar(vx_graph graph, vx_enum data_type)
{
AgoData * data = NULL;
if (agoIsValidGraph(graph)) {
CAgoLock lock(graph->cs);
const char * desc_type = agoEnum2Name(data_type);
if (data_type && !desc_type) {
desc_type = agoGetUserStructName(graph->ref.context, data_type);
}
if(!data_type || desc_type) {
char desc[512];
if (desc_type) sprintf(desc, "scalar-virtual:%s,0", desc_type);
else sprintf(desc, "scalar-virtual:0,0");
data = agoCreateDataFromDescription(graph->ref.context, graph, desc, true);
if (data) {
agoGenerateVirtualDataName(graph, "scalar", data->name);
agoAddData(&graph->dataList, data);
}
}
}
return (vx_scalar)data;
}
/*! \brief Creates a reference to a scalar object. Also see \ref sub_node_parameters.
* \param [in] context The reference to the system context.
* \param [in] data_type The type of data to hold. Must be greater than
* \ref VX_TYPE_INVALID and less than or equal to \ref VX_TYPE_VENDOR_STRUCT_END.
* Or must be a \ref vx_enum returned from \ref vxRegisterUserStruct.
* \param [in] ptr The pointer to the initial value of the scalar.
* \param [in] size Size of data at ptr in bytes.
* \ingroup group_scalar
* \returns A scalar reference \ref vx_scalar. Any possible errors preventing a
* successful creation should be checked using \ref vxGetStatus.
*/
VX_API_ENTRY vx_scalar VX_API_CALL vxCreateScalarWithSize(vx_context context, vx_enum data_type, const void *ptr, vx_size size)
{
AgoData * data = NULL;
if (agoIsValidContext(context)) {
CAgoLock lock(context->cs);
vx_size data_size = 0;
if(data_type != VX_TYPE_STRING_AMD) {
data_size = agoType2Size(context, data_type);
if(data_size == 0 || data_size != size) {
return NULL;
}
}
data = agoCreateDataFromDescription(context, NULL, "scalar:UINT32,0", true);
if (data) {
agoAddData(&context->dataList, data);
data->u.scalar.type = data_type;
data->u.scalar.itemsize = size;
switch (data_type) {
case VX_TYPE_ENUM:
if (ptr) data->u.scalar.u.e = *(vx_enum *)ptr;
break;
case VX_TYPE_UINT32:
if (ptr) data->u.scalar.u.u = *(vx_uint32 *)ptr;
break;
case VX_TYPE_INT32:
if (ptr) data->u.scalar.u.i = *(vx_int32 *)ptr;
break;
case VX_TYPE_UINT16:
if (ptr) data->u.scalar.u.u = *(vx_uint16 *)ptr;
break;
case VX_TYPE_INT16:
if (ptr) data->u.scalar.u.i = *(vx_int16 *)ptr;
break;
case VX_TYPE_UINT8:
if (ptr) data->u.scalar.u.u = *(vx_uint8 *)ptr;
break;
case VX_TYPE_INT8:
if (ptr) data->u.scalar.u.i = *(vx_int8 *)ptr;
break;
case VX_TYPE_CHAR:
if (ptr) data->u.scalar.u.i = *(vx_char *)ptr;
break;
case VX_TYPE_FLOAT32:
if (ptr) data->u.scalar.u.f = *(vx_float32 *)ptr;
break;
case VX_TYPE_FLOAT16:
if (ptr) data->u.scalar.u.u = *(vx_uint16 *)ptr;
break;
case VX_TYPE_SIZE:
if (ptr) data->u.scalar.u.s = *(vx_size *)ptr;
break;
case VX_TYPE_BOOL:
if (ptr) data->u.scalar.u.u = *(vx_bool *)ptr;
break;
case VX_TYPE_DF_IMAGE:
if (ptr) data->u.scalar.u.df = *(vx_df_image *)ptr;
break;
case VX_TYPE_FLOAT64:
if (ptr) data->u.scalar.u.f64 = *(vx_float64 *)ptr;
break;
case VX_TYPE_INT64:
if (ptr) data->u.scalar.u.i64 = *(vx_int64 *)ptr;
break;
case VX_TYPE_UINT64:
if (ptr) data->u.scalar.u.u64 = *(vx_uint64 *)ptr;
break;
default:
if(data_type == VX_TYPE_STRING_AMD) {
data->u.scalar.itemsize = sizeof(char *);
data->size = VX_MAX_STRING_BUFFER_SIZE_AMD;
}
else {
data->u.scalar.itemsize = data->size = data_size;
}
data->buffer_allocated = data->buffer = (vx_uint8 *)agoAllocMemory(data->size);
if (data->buffer) {
memset(data->buffer, 0, data->size);
if (ptr) {
if(data_type == VX_TYPE_STRING_AMD) {
strncpy((char *)data->buffer, (const char *)ptr, VX_MAX_STRING_BUFFER_SIZE_AMD);
data->buffer[VX_MAX_STRING_BUFFER_SIZE_AMD - 1] = 0; // NUL terminate string in case of overflow
}
else {
memcpy(data->buffer, ptr, size);
}
}
data->isInitialized = vx_true_e;
}
else {
agoReleaseData(data, true);
data = NULL;
}
break;
}
}
}
return (vx_scalar)data;
}
/*! \brief Releases a reference to a scalar object.
* The object may not be garbage collected until its total reference count is zero.
* \param [in] scalar The pointer to the scalar to release.
* \ingroup group_scalar
* \post After returning from this function the reference is zeroed.
* \return A \ref vx_status_e enumeration.
* \retval VX_SUCCESS No errors.
* \retval VX_ERROR_INVALID_REFERENCE If graph is not a \ref vx_graph.
*/
VX_API_ENTRY vx_status VX_API_CALL vxReleaseScalar(vx_scalar *scalar)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (scalar && agoIsValidData((AgoData *)*scalar, VX_TYPE_SCALAR)) {
if (!agoReleaseData((AgoData *)*scalar, true)) {
*scalar = NULL;
status = VX_SUCCESS;
}
}
return status;
}
/*! \brief Queries attributes from a scalar.
* \param [in] scalar The scalar object.
* \param [in] attribute The enumeration to query. Use a \ref vx_scalar_attribute_e enumeration.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \ingroup group_scalar
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryScalar(vx_scalar scalar, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
AgoData * data = (AgoData *)scalar;
if (agoIsValidData(data, VX_TYPE_SCALAR)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_SCALAR_ATTRIBUTE_TYPE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = data->u.scalar.type;
status = VX_SUCCESS;
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Gets the scalar value out of a reference.
* \note Use this in conjunction with Query APIs that return references which
* should be converted into values.
* \ingroup group_scalar
* \param [in] ref The reference from which to get the scalar value.
* \param [out] ptr An appropriate typed pointer that points to a location to which to copy
* the scalar value.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_INVALID_REFERENCE If the ref is not a valid
* reference.
* \retval VX_ERROR_INVALID_PARAMETERS If \a ptr is NULL.
* \retval VX_ERROR_INVALID_TYPE If the type does not match the type in the reference or is a bad value.
*/
VX_API_ENTRY vx_status VX_API_CALL vxReadScalarValue(vx_scalar ref, void *ptr)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
AgoData * data = (AgoData *)ref;
if (agoIsValidData(data, VX_TYPE_SCALAR)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
status = VX_SUCCESS;
switch (data->u.scalar.type)
{
case VX_TYPE_ENUM:
*(vx_enum *)ptr = data->u.scalar.u.e;
break;
case VX_TYPE_UINT32:
*(vx_uint32 *)ptr = data->u.scalar.u.u;
break;
case VX_TYPE_INT32:
*(vx_int32 *)ptr = data->u.scalar.u.i;
break;
case VX_TYPE_UINT16:
*(vx_uint16 *)ptr = data->u.scalar.u.u;
break;
case VX_TYPE_INT16:
*(vx_int16 *)ptr = data->u.scalar.u.i;
break;
case VX_TYPE_UINT8:
*(vx_uint8 *)ptr = data->u.scalar.u.u;
break;
case VX_TYPE_INT8:
*(vx_int8 *)ptr = data->u.scalar.u.i;
break;
case VX_TYPE_CHAR:
*(vx_char *)ptr = data->u.scalar.u.i;
break;
case VX_TYPE_FLOAT32:
*(vx_float32 *)ptr = data->u.scalar.u.f;
break;
case VX_TYPE_SIZE:
*(vx_size *)ptr = data->u.scalar.u.s;
break;
case VX_TYPE_BOOL:
*(vx_bool *)ptr = data->u.scalar.u.u ? vx_true_e : vx_false_e;
break;
case VX_TYPE_DF_IMAGE:
*(vx_df_image *)ptr = data->u.scalar.u.df;
break;
case VX_TYPE_FLOAT64:
*(vx_float64 *)ptr = data->u.scalar.u.f64;
break;
case VX_TYPE_UINT64:
*(vx_uint64 *)ptr = data->u.scalar.u.u64;
break;
case VX_TYPE_INT64:
*(vx_int64 *)ptr = data->u.scalar.u.i64;
break;
case VX_TYPE_STRING_AMD:
strcpy((char *)ptr, (const char *)data->buffer);
break;
default:
if (data->buffer) {
memcpy(ptr, data->buffer, data->size);
break;
}
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Sets the scalar value in a reference.
* \note Use this in conjunction with Parameter APIs that return references
* to parameters that need to be altered.
* \ingroup group_scalar
* \param [in] ref The reference from which to get the scalar value.
* \param [in] ptr An appropriately typed pointer that points to a location to which to copy
* the scalar value.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_INVALID_REFERENCE If the ref is not a valid
* reference.
* \retval VX_ERROR_INVALID_PARAMETERS If \a ptr is NULL.
* \retval VX_ERROR_INVALID_TYPE If the type does not match the type in the reference or is a bad value.
*/
VX_API_ENTRY vx_status VX_API_CALL vxWriteScalarValue(vx_scalar ref, const void *ptr)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
AgoData * data = (AgoData *)ref;
if (agoIsValidData(data, VX_TYPE_SCALAR) && !data->isVirtual) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
// TBD: need sem-lock for thread safety
status = VX_SUCCESS;
switch (data->u.scalar.type)
{
case VX_TYPE_ENUM:
data->u.scalar.u.e = *(vx_enum *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_UINT32:
data->u.scalar.u.u = *(vx_uint32 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_INT32:
data->u.scalar.u.i = *(vx_int32 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_UINT16:
data->u.scalar.u.u = *(vx_uint16 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_INT16:
data->u.scalar.u.i = *(vx_int16 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_UINT8:
data->u.scalar.u.u = *(vx_uint8 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_INT8:
data->u.scalar.u.i = *(vx_int8 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_CHAR:
data->u.scalar.u.i = *(vx_char *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_FLOAT32:
data->u.scalar.u.f = *(vx_float32 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_SIZE:
data->u.scalar.u.s = *(vx_size *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_BOOL:
data->u.scalar.u.u = *(vx_bool *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_DF_IMAGE:
data->u.scalar.u.df = *(vx_df_image *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_FLOAT64:
data->u.scalar.u.f64 = *(vx_float64 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_UINT64:
data->u.scalar.u.u64 = *(vx_uint64 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_INT64:
data->u.scalar.u.i64 = *(vx_int64 *)ptr;
data->isInitialized = vx_true_e;
break;
case VX_TYPE_STRING_AMD:
strncpy((char *)data->buffer, (const char *)ptr, VX_MAX_STRING_BUFFER_SIZE_AMD);
data->buffer[VX_MAX_STRING_BUFFER_SIZE_AMD - 1] = 0; // NUL terminate string in case of overflow
data->isInitialized = vx_true_e;
break;
default:
if (ptr) {
memcpy(data->buffer,ptr, data->size);
break;
}
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Allows the application to copy from/into a scalar object.
* \param [in] scalar The reference to the scalar object that is the source or the
* destination of the copy.
* \param [in] user_ptr The address of the memory location where to store the requested data
* if the copy was requested in read mode, or from where to get the data to store into the
* scalar object if the copy was requested in write mode. In the user memory, the scalar is
* a variable of the type corresponding to \ref VX_SCALAR_TYPE.
* The accessible memory must be large enough to contain this variable.
* \param [in] usage This declares the effect of the copy with regard to the scalar object
* using the \ref vx_accessor_e enumeration. Only VX_READ_ONLY and VX_WRITE_ONLY
* are supported:
* \arg VX_READ_ONLY means that data are copied from the scalar object into the user memory.
* \arg VX_WRITE_ONLY means that data are copied into the scalar object from the user memory.
* \param [in] user_mem_type A \ref vx_memory_type_e enumeration that specifies
* the memory type of the memory referenced by the user_addr.
* \return A \ref vx_status_e enumeration.
* \retval VX_ERROR_INVALID_REFERENCE The scalar reference is not actually a scalar reference.
* \retval VX_ERROR_INVALID_PARAMETERS An other parameter is incorrect.
* \ingroup group_scalar
*/
VX_API_ENTRY vx_status VX_API_CALL vxCopyScalar(vx_scalar scalar_, void *user_ptr, vx_enum usage, vx_enum user_mem_type)
{
AgoData * scalar = (AgoData *)scalar_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(scalar, VX_TYPE_SCALAR))
{
status = VX_ERROR_INVALID_PARAMETERS;
if ((user_mem_type == VX_MEMORY_TYPE_HOST) && user_ptr) {
if (usage == VX_READ_ONLY)
status = vxReadScalarValue(scalar_, user_ptr);
else if (usage == VX_WRITE_ONLY)
status = vxWriteScalarValue(scalar_, user_ptr);
}
}
return status;
}
VX_API_ENTRY vx_status VX_API_CALL vxCopyScalarWithSize(vx_scalar scalar_, vx_size size, void *user_ptr, vx_enum usage, vx_enum user_mem_type)
{
AgoData * scalar = (AgoData *)scalar_;
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidData(scalar, VX_TYPE_SCALAR))
{
status = VX_ERROR_INVALID_PARAMETERS;
if ((user_mem_type == VX_MEMORY_TYPE_HOST) && user_ptr && (scalar->u.scalar.itemsize == size)) {
if (usage == VX_READ_ONLY)
status = vxReadScalarValue(scalar_, user_ptr);
else if (usage == VX_WRITE_ONLY)
status = vxWriteScalarValue(scalar_, user_ptr);
}
}
return status;
}
/*==============================================================================
REFERENCE
=============================================================================*/
/*! \brief Queries any reference type for some basic information (count, type).
* \param [in] ref The reference to query.
* \param [in] attribute The value for which to query. Use \ref vx_reference_attribute_e.
* \param [out] ptr The location at which to store the resulting value.
* \param [in] size The size of the container to which \a ptr points.
* \return A \ref vx_status_e enumeration.
* \ingroup group_reference
*/
VX_API_ENTRY vx_status VX_API_CALL vxQueryReference(vx_reference ref, vx_enum attribute, void *ptr, vx_size size)
{
vx_status status = VX_ERROR_INVALID_REFERENCE;
if (agoIsValidReference(ref)) {
status = VX_ERROR_INVALID_PARAMETERS;
if (ptr) {
switch (attribute)
{
case VX_REFERENCE_COUNT:
if (size == sizeof(vx_uint32)) {
*(vx_uint32 *)ptr = ref->external_count;
status = VX_SUCCESS;
}
break;
case VX_REFERENCE_TYPE:
if (size == sizeof(vx_enum)) {
*(vx_enum *)ptr = ref->type;
status = VX_SUCCESS;
}
break;
case VX_REFERENCE_NAME:
if (size == sizeof(vx_char*)) {
if(ref->type == VX_TYPE_GRAPH)
{
AgoGraph * graph = (AgoGraph *)ref;
//strncpy((char *)ptr, data->name.c_str(), size);
*(vx_char**)ptr = &graph->name[0];
status = VX_SUCCESS;
}
else
{
AgoData * data = (AgoData *)ref;
//strncpy((char *)ptr, data->name.c_str(), size);
*(vx_char**)ptr = &data->name[0];
status = VX_SUCCESS;
}
}
break;
default:
status = VX_ERROR_NOT_SUPPORTED;
break;
}
}
}
return status;
}
/*! \brief Releases a reference. The reference may potentially refer to multiple OpenVX objects of different types.
* This function can be used instead of calling a specific release function for each individual object type
* (e.g. vxRelease