<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "gl_context.h"

static const EGLint _display_attribs[] =
    {
        EGL_RED_SIZE, 8,
        EGL_GREEN_SIZE, 8,
        EGL_BLUE_SIZE, 8,
        EGL_ALPHA_SIZE, 8,
        EGL_NONE
    };

static const EGLint _context_attribs[] =
    {
        EGL_CONTEXT_CLIENT_VERSION, 3,
        EGL_CONTEXT_MAJOR_VERSION_KHR, 3,
        EGL_CONTEXT_MINOR_VERSION_KHR, 1,
        EGL_NONE
    };

static EGLContext _context = EGL_NO_CONTEXT;
static EGLSurface _surface = EGL_NO_SURFACE;
static EGLDisplay _display = EGL_NO_DISPLAY;

// -------------------------------------------------------------------------------------------------

void gl_context_create(struct ANativeWindow *window)
{
    EGLint num_configs;
    EGLConfig *configs;
    EGLint r, g, b;
    EGLConfig selected_config;

    size_t i;

    if (window == NULL)
        return;

    _display = eglGetDisplay(EGL_DEFAULT_DISPLAY);
    eglInitialize(_display, 0, 0);

    // Get a config
    eglChooseConfig(_display, _display_attribs, NULL, 0, &num_configs);
    configs = (EGLConfig *) malloc(num_configs * sizeof(EGLConfig));

    eglChooseConfig(_display, _display_attribs, configs, num_configs, &num_configs);

    for (i = 0; i < num_configs; i++)
    {
        if (eglGetConfigAttrib(_display, configs[i], EGL_RED_SIZE, &r)
            && eglGetConfigAttrib(_display, configs[i], EGL_GREEN_SIZE, &g)
            && eglGetConfigAttrib(_display, configs[i], EGL_BLUE_SIZE, &b)
            && r == 8 && g == 8 && b == 8)
        {
            selected_config = configs[i];
            break;
        }
    }

    if (i == num_configs)
    {
        LOGE("Default config loaded");
        selected_config = configs[0];
    }

    free(configs);

    EGLint format;
    eglGetConfigAttrib(_display, selected_config, EGL_NATIVE_VISUAL_ID, &format);

    _surface = eglCreateWindowSurface
        (
            _display,
            selected_config,
            window,
            NULL
        );
    _context = eglCreateContext(_display, selected_config, NULL, _context_attribs);

    // Load OpenGl functions
    gl_functions_load();

    if (eglMakeCurrent(_display, _surface, _surface, _context) ==
        EGL_FALSE)
    {
        LOGE("Unable to make current");
    }

    // Check init OK
    GLenum infos[] = {GL_VENDOR, GL_RENDERER, GL_VERSION, GL_EXTENSIONS};
    for (i = 0; i < 4; i++)
    {
        const GLubyte *info = glGetString(infos[i]);
        LOGD("OpenGL info : %s", info);
    }

    eglMakeCurrent(_display, EGL_NO_SURFACE, EGL_NO_SURFACE, EGL_NO_CONTEXT);
}

void gl_context_init_computing(struct compute_data *data)
{
    gl_computing_init(data);
}

void gl_context_update_computing(struct compute_data *data)
{
    gl_computing_update(data);
}

void gl_context_perform_computing(struct compute_data *data)
{
    gl_computing_compute(data);
}

void gl_context_destroy_computing(struct compute_data *data)
{
    gl_computing_destroy(data);
}

void gl_context_init_frame(struct display_data *data)
{
    rendering_init(data);
}

void gl_context_update_frame(struct display_data *data)
{
    rendering_update(data);
}

void gl_context_clear_frame(struct display_data *data)
{
    rendering_clear(data);
}

void gl_context_resize_frame(struct display_data *data)
{
    GLsizei w;
    GLsizei h;

    eglQuerySurface(_display, _surface, EGL_WIDTH, &(w));
    eglQuerySurface(_display, _surface, EGL_HEIGHT, &(h));

    rendering_resize(data, w, h);
}

void gl_context_draw_frame(struct display_data *data)
{
    rendering_draw(data);
    eglSwapBuffers(_display, _surface);
}

void gl_context_destroy_frame(struct display_data *data)
{
    rendering_destroy(data);
}

void gl_context_bind()
{
    if (eglGetCurrentContext() != _context)
    {
        if (eglMakeCurrent(_display, _surface, _surface, _context) == EGL_FALSE)
        {
            LOGE("Unable to bind the current EGL context");
        }
    }
}

void gl_context_unbind()
{
    eglMakeCurrent(_display, EGL_NO_SURFACE, EGL_NO_SURFACE, EGL_NO_CONTEXT);
}

void gl_context_destroy()
{
    if (_display != EGL_NO_DISPLAY)
    {
        if (eglMakeCurrent(_display, _surface, _surface, _context) ==
            EGL_FALSE)
        {
            LOGE("Unable to make current for destroying");
            return;
        }

        if (_context != EGL_NO_CONTEXT)
        {
            eglDestroyContext(_display, _context);
        }

        if (_surface != EGL_NO_SURFACE)
        {
            eglDestroySurface(_display, _surface);
        }

        eglTerminate(_display);
    }

    _display = EGL_NO_DISPLAY;
    _context = EGL_NO_CONTEXT;
    _surface = EGL_NO_SURFACE;
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "global.h"

void checkCLError(const char *message, const cl_int error_id)
{
    switch (error_id)
    {
        case CL_SUCCESS:
            LOGD("%s : CL_SUCCESS", message);
            break;
        case CL_INVALID_DEVICE:
            LOGE("%s : CL_INVALID_DEVICE", message);
            break;
        case CL_INVALID_VALUE:
            LOGE("%s : CL_INVALID_VALUE", message);
            break;
        case CL_INVALID_GL_SHAREGROUP_REFERENCE_KHR:
            LOGE("%s : CL_INVALID_GL_SHAREGROUP_REFERENCE_KHR", message);
            break;
        case CL_INVALID_OPERATION:
            LOGE("%s : CL_INVALID_OPERATION", message);
            break;
        case CL_OUT_OF_RESOURCES:
            LOGE("%s : CL_OUT_OF_RESOURCES", message);
            break;
        case CL_INVALID_CONTEXT:
            LOGE("%s : CL_INVALID_CONTEXT", message);
            break;
        case CL_OUT_OF_HOST_MEMORY:
            LOGE("%s : CL_OUT_OF_HOST_MEMORY", message);
            break;
        case CL_INVALID_PROGRAM:
            LOGE("%s : CL_INVALID_PROGRAM", message);
            break;
        case CL_INVALID_BINARY:
            LOGE("%s : CL_INVALID_BINARY", message);
            break;
        case CL_INVALID_BUILD_OPTIONS:
            LOGE("%s : CL_INVALID_BUILD_OPTIONS", message);
            break;
        case CL_BUILD_PROGRAM_FAILURE:
            LOGE("%s : CL_BUILD_PROGRAM_FAILURE", message);
            break;
        case CL_COMPILER_NOT_AVAILABLE:
            LOGE("%s : CL_COMPILER_NOT_AVAILABLE", message);
            break;
        case CL_INVALID_BUFFER_SIZE:
            LOGE("%s : CL_INVALID_BUFFER_SIZE", message);
            break;
        case CL_MEM_OBJECT_ALLOCATION_FAILURE:
            LOGE("%s : CL_MEM_OBJECT_ALLOCATION_FAILURE", message);
            break;
        case CL_INVALID_IMAGE_FORMAT_DESCRIPTOR:
            LOGE("%s : CL_INVALID_IMAGE_FORMAT_DESCRIPTOR", message);
            break;
        case CL_INVALID_IMAGE_DESCRIPTOR:
            LOGE("%s : CL_INVALID_IMAGE_DESCRIPTOR", message);
            break;
        case CL_INVALID_IMAGE_SIZE:
            LOGE("%s : CL_INVALID_IMAGE_SIZE", message);
            break;
        case CL_INVALID_HOST_PTR:
            LOGE("%s : CL_INVALID_HOST_PTR", message);
            break;
        case CL_IMAGE_FORMAT_NOT_SUPPORTED:
            LOGE("%s : CL_IMAGE_FORMAT_NOT_SUPPORTED", message);
            break;
        default:
            LOGE("%s : unknown error", message);
            break;
    }
}

void checkGLError(const char *message)
{
    GLenum err = glGetError();
    switch (err)
    {
        case
            GL_NO_ERROR:
            LOGI("%s OK", message);
            break;
        case
            GL_INVALID_ENUM:
            LOGE("%s GL error : INVALID ENUM", message);
            break;
        case
            GL_INVALID_VALUE:
            LOGE("%s GL error : INVALID VALUE", message);
            break;
        case
            GL_INVALID_OPERATION:
            LOGE("%s GL error : INVALID OPERATION", message);
            break;
        case
            GL_INVALID_FRAMEBUFFER_OPERATION:
            LOGE("%s GL error : INVALID FRAMEBUFFER OPERATION", message);
            break;
        case
            GL_OUT_OF_MEMORY:
            LOGE("%s GL error : OUT OF MEMORY", message);
            break;
    }
}

void checkFrameBufferCompleteness(const char *message)
{
    GLenum err = glCheckFramebufferStatus(GL_FRAMEBUFFER);
    switch (err)
    {
        case GL_FRAMEBUFFER_INCOMPLETE_ATTACHMENT:
            LOGE("%s framebuffer incomplete : GL_FRAMEBUFFER_INCOMPLETE_ATTACHMENT", message);
            break;
#ifdef __ANDROID
        case GL_FRAMEBUFFER_INCOMPLETE_DIMENSIONS:
            LOGE("%s framebuffer incomplete : GL_FRAMEBUFFER_INCOMPLETE_DIMENSIONS", message);
            break;
#endif
        case GL_FRAMEBUFFER_INCOMPLETE_MISSING_ATTACHMENT:
            LOGE("%s framebuffer incomplete : GL_FRAMEBUFFER_INCOMPLETE_MISSING_ATTACHMENT",
                 message);
            break;
        case GL_FRAMEBUFFER_UNSUPPORTED:
            LOGE("%s framebuffer incomplete : GL_FRAMEBUFFER_UNSUPPORTED", message);
            break;
        case GL_FRAMEBUFFER_COMPLETE:
            LOGD("%s framebuffer complete", message);
            break;
    }
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_GL_THREAD_H
#define OPENINGS_GL_THREAD_H

#include <pthread.h>
#include <unistd.h>
#include <stdlib.h>
#include <stdint.h>
#include <string.h>
#include <errno.h>
#include <omp.h>
#include <EGL/egl.h>
#include <EGL/eglext.h>
#include <android/native_window.h>

#include "cl_context/cl_context.h"
#include "global/global.h"
#include "gl/gl_context.h"
#include "model/model.h"
#include "rendering/rendering.h"
#include "gl_computing/gl_computing.h"

#define GL_THREAD_SLEEP_TIME                10000       // 10 ms

#define GL_THREAD_RENDER_CLEAR_REQUEST      0b00000001
#define GL_THREAD_RENDER_UPDATE_REQUEST     0b00000010
#define GL_THREAD_RENDER_RESIZE_REQUEST     0b00000100
#define GL_THREAD_COMPUTE_UPDATE_REQUEST    0b00001000
#define GL_THREAD_COMPUTE_PERFORM_REQUEST   0b00010000

__BEGIN_DECLS

extern void gl_thread_set_flag(uint8_t flag);

extern void gl_thread_start(struct ANativeWindow *window);

extern void gl_thread_stop();

__END_DECLS

#endif // OPENINGS_GL_THREAD_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_RENDERING_H
#define OPENINGS_RENDERING_H

#include <stdint.h>
#include <stdlib.h>
#include <string.h>
#include <GLES3/gl3.h>

#include "global/global.h"
#include "gl/gl_31_functions.h"
#include "gl/shader.h"
#include "model/model.h"

__BEGIN_DECLS

extern void rendering_init(struct display_data *data);

extern void rendering_resize(struct display_data *data, GLsizei w, GLsizei h);

extern void rendering_update(struct display_data *data);

extern void rendering_clear(struct display_data *data);

extern void rendering_draw(struct display_data *data);

extern void rendering_destroy(struct display_data *data);

__END_DECLS

#endif // OPENINGS_RENDERING_H
<file_sep># Created by <NAME>
#
# This program is free software: you can redistribute it and/or modify
# it under the terms of the GNU General Public License as published by
# the Free Software Foundation, either version 3 of the License, or
# (at your option) any later version.
#
# This program is distributed in the hope that it will be useful,
# but WITHOUT ANY WARRANTY; without even the implied warranty of
# MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
# GNU General Public License for more details.
#
# You should have received a copy of the GNU General Public License

LOCAL_PATH := $(call my-dir)

include $(CLEAR_VARS)

LOCAL_ARM_NEON := true

ifeq ($(TARGET_ARCH_ABI),armeabi-v7a)
LOCAL_CFLAGS := -mfloat-abi=softfp -mfpu=neon
endif

LOCAL_MODULE := native_interface

LOCAL_C_INCLUDES += $(LOCAL_PATH)/ \
                        $(LOCAL_PATH)/application/ \
                        $(LOCAL_PATH)/CL/ \
                        $(LOCAL_PATH)/computing/ \
                        $(LOCAL_PATH)/gl/ \
                        $(LOCAL_PATH)/global/ \
                        $(LOCAL_PATH)/model/ \
                        $(LOCAL_PATH)/rendering/ \
                        $(LOCAL_PATH)/tango/

LOCAL_SRC_FILES := $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/application/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/computing/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/gl/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/global/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/model/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/rendering/*.c)) \
                    $(subst $(LOCAL_PATH)/,,$(wildcard $(LOCAL_PATH)/tango/*.c))

LOCAL_LDLIBS    :=  -lGLESv3 \
                    -lEGL \
                    -llog \
                    -lc \
                    -lm \
                    -landroid \
                    -fopenmp \
                    -lOpenCL \
                    -L$(LOCAL_PATH)/../JNILibs

include $(BUILD_SHARED_LIBRARY)<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_GLOBAL_H
#define OPENINGS_GLOBAL_H

#ifndef _REENTRANT
# define _REENTRANT
#endif

#ifndef CL_TARGET_OPENCL_VERSION
#   define CL_TARGET_OPENCL_VERSION    220
#endif

#include <CL/opencl.h>

#include <pthread.h>
#include <stdint.h>
#include <stdlib.h>
#include <string.h>
#include <unistd.h>

#include <android/log.h>
#include <GLES3/gl3.h>

#include "CL/opencl.h"

// ---------------------------------------------------------------------------------------------

#define LOG_TAG                     "native-activity"

/**
 * Different constants related to the depth and color camera (width, height, resolutions, ...)
 */
#define COLOR_CAMERA_WIDTH          1920
#define COLOR_CAMERA_HEIGHT         1080
#define COLOR_CAMERA_RESOLUTION     (COLOR_CAMERA_WIDTH * COLOR_CAMERA_HEIGHT)
#define COLOR_CAMERA_IMAGE_SIZE     (COLOR_CAMERA_RESOLUTION * 3)

/**
 * OpenGL offsets
 */
#define BUFFER_OFFSET(offset)       ((char*)NULL + (offset))
#define BUFFER_SIZE(size)           ((GLsizeiptr)((char*)NULL + size))

/**
 * Useful functions
 */
#define NOT_ZERO(P)                 (((P)[0] + (P)[1] + (P)[2]) != 0.0f)
#define CLAMP(x, a, b)              (x < a ? a : (x > b ? b : x))
#define SWAP_POINTERS(p1, p2)       swap_pointers((void**)p1, (void**)p2)

/**
 * Helper functions to generate an Android log
 */
# define LOGI(...)  ((void)__android_log_print(ANDROID_LOG_INFO, LOG_TAG, __VA_ARGS__))
# define LOGW(...)  ((void)__android_log_print(ANDROID_LOG_WARN, LOG_TAG, __VA_ARGS__))
# define LOGD(...)  ((void)__android_log_print(ANDROID_LOG_DEBUG, LOG_TAG, __VA_ARGS__))
# define LOGE(...)  ((void)__android_log_print(ANDROID_LOG_ERROR, LOG_TAG, __VA_ARGS__))

#define DEBUG_GL
#ifdef DEBUG_GL
#define __glcall__(P) P; checkGLError(#P)
#else
#define __glcall__(P)  P
#endif


// ---------------------------------------------------------------------------------------------

__BEGIN_DECLS

/**
 *
 * @param message
 * @param error_id
 */
extern void checkCLError(const char *message, const cl_int error_id);

/**
 * Check the last OpenGL error and displays the associated message
 * @param message Any message we want to associate with the error checking.
 */
extern void checkGLError(const char *message);

/**
 * Check the last OpenGl framebuffer status and displays the associated message.
 * @param message Any message we want to associate with the framebuffer checking.
 */
extern void checkFrameBufferCompleteness(const char *message);

/**
 * Perform a swap between two pointers . The swap is performed atomically so it can be used
 * to swap data between different threads.
 * @param p1
 * @param p2
 */
static inline void swap_pointers(void **p1, void **p2)
{
    void *tmp; // = *p1;
#pragma omp atomic read
    tmp = *p1;

#pragma omp atomic capture
    {
        *p1 = *p2;
        *p2 = tmp;
    }
}

__END_DECLS

#endif // OPENINGS_GLOBAL_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_GL_CONTEXT_H
#define OPENINGS_GL_CONTEXT_H

#include <stdint.h>
#include <stdlib.h>

#include <android/native_window.h>
#include <EGL/egl.h>
#include <EGL/eglext.h>

#include "gl_computing/gl_computing.h"
#include "gl/gl_31_functions.h"
#include "global/global.h"
#include "model/model.h"
#include "rendering/rendering.h"

__BEGIN_DECLS

extern void gl_context_create(struct ANativeWindow *window);

extern void gl_context_bind();

extern void gl_context_unbind();

#if (USE_OPENCL == 0)

extern void gl_context_init_computing(struct compute_data *data);

extern void gl_context_update_computing(struct compute_data *data);

extern void gl_context_perform_computing(struct compute_data *data);

extern void gl_context_destroy_computing(struct compute_data *data);

#endif

extern void gl_context_init_frame(struct display_data *data);

extern void gl_context_update_frame(struct display_data *data);

extern void gl_context_clear_frame(struct display_data *data);

extern void gl_context_resize_frame(struct display_data *data);

extern void gl_context_draw_frame(struct display_data *data);

extern void gl_context_destroy_frame(struct display_data *data);

extern void gl_context_destroy();

__END_DECLS

#endif // OPENINGS_GL_CONTEXT_H
<file_sep>/*
 * Copyright (c) 2018.
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 * along with this program.  If not, see <http://www.gnu.org/licenses/>.
 */

/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 * along with this program.  If not, see <http://www.gnu.org/licenses/>.
 */

package limsi.fr.openings;

import android.content.Context;
import android.hardware.Camera;
import android.util.Log;
import android.view.SurfaceHolder;
import android.view.SurfaceView;

import java.io.IOException;

public class CameraSurfaceView extends SurfaceView
        implements SurfaceHolder.Callback, Camera.PreviewCallback {

    private Camera _camera;
    private SurfaceHolder _holder;

    public CameraSurfaceView(Context context, Camera camera) {
        super(context);
        _camera = camera;
        _holder = getHolder();
        _holder.addCallback(this);
    }

    @Override
    public void onPreviewFrame(byte[] img, Camera camera) {
        nativeSetImage(img, camera.getParameters().getPreviewSize().width,
                camera.getParameters().getPreviewSize().height);
    }

    @Override
    public void surfaceCreated(SurfaceHolder holder) {
        Log.d("CameraSurfaceView", "Surface created");
    }

    @Override
    public void surfaceChanged(SurfaceHolder holder, int format, int width, int height) {
        Log.d("CameraSurfaceView", "Surface changed");

        if (holder.getSurface() == null) {
            return;
        }

        try {
            _camera.stopPreview();
        } catch (Exception e) {

        }

        try {
            _camera.setPreviewDisplay(holder);
            _camera.setPreviewCallback(this);
            _camera.startPreview();
        } catch (IOException e) {
            Log.e("CameraSurfaceView", e.getMessage());
        }
    }

    @Override
    public void surfaceDestroyed(SurfaceHolder holder) {

    }

    private static native void nativeSetImage(byte img[], int width, int height);
}
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "cl_grayscale.h"

static struct compute_info _info;
static GLuint _output_texture_id;

static float _input_image[3][COLOR_CAMERA_RESOLUTION];
static float _output_image[4 * COLOR_CAMERA_RESOLUTION];

static cl_program _program;

static cl_mem _input_red_image_id;
static cl_mem _input_green_image_id;
static cl_mem _input_blue_image_id;
static cl_mem _output_image_id;

static cl_image_format _input_format =
    {
        .image_channel_order = CL_R,
        .image_channel_data_type = CL_FLOAT
    };
static cl_image_desc _input_description =
    {
        .image_type     =   CL_MEM_OBJECT_IMAGE2D,
        .image_width = COLOR_CAMERA_WIDTH,
        .image_height = COLOR_CAMERA_HEIGHT,
        .image_array_size = 0,
        .image_row_pitch = 0,
        .image_slice_pitch = 0,
        .num_mip_levels = 0,
        .num_samples = 0,
        .buffer = NULL
    };
static cl_image_format _output_format =
    {
        .image_channel_order = CL_RGBA,
        .image_channel_data_type = CL_FLOAT
    };
static cl_image_desc _output_description =
    {
        .image_type = CL_MEM_OBJECT_IMAGE2D,
        .image_width = COLOR_CAMERA_WIDTH,
        .image_height = COLOR_CAMERA_HEIGHT,
        .image_array_size = 0,
        .image_row_pitch = 0,
        .image_slice_pitch = 0,
        .num_mip_levels = 0,
        .num_samples = 0,
        .buffer = NULL
    };

void grayscale_init(struct compute_info *info, const GLuint output_texture)
{
    memset(_input_image[0], 0, COLOR_CAMERA_RESOLUTION * sizeof(float));
    memset(_input_image[1], 0, COLOR_CAMERA_RESOLUTION * sizeof(float));
    memset(_input_image[2], 0, COLOR_CAMERA_RESOLUTION * sizeof(float));

    memcpy(&_info, info, sizeof(struct compute_info));
    _output_texture_id = output_texture;

    _program = kernel_create_program(_info.context, &(_info.device_id), grayscale_kernel_source);
    _input_red_image_id = clCreateImage
        (
            _info.context,
            CL_MEM_READ_ONLY,
            &_input_format,
            &_input_description,
            NULL,
            NULL
        );
    _input_green_image_id = clCreateImage
        (
            _info.context,
            CL_MEM_READ_ONLY,
            &_input_format,
            &_input_description,
            NULL,
            NULL
        );
    _input_blue_image_id = clCreateImage
        (
            _info.context,
            CL_MEM_READ_ONLY,
            &_input_format,
            &_input_description,
            NULL,
            NULL
        );

    _output_image_id = clCreateFromGLTexture
        (
            _info.context,
            CL_MEM_WRITE_ONLY,
            GL_TEXTURE_2D,
            0, _output_texture_id, NULL
        );
}

void grayscale_set_input_image(const float *img)
{
    size_t i;
    size_t index = 0;
    for (i = 0; i < COLOR_CAMERA_RESOLUTION; i++)
    {
        _input_image[0][i] = img[index++];
        _input_image[1][i] = img[index++];
        _input_image[2][i] = img[index++];
    }

    // Update CL objects
    cl_int err;
    size_t img_index[3] = {0, 0, 0};
    size_t img_region[3] = {COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT, 1};

    err = clEnqueueWriteImage
        (
            _info.command_queue,
            _input_red_image_id,
            CL_FALSE,
            img_index, img_region,
            0, 0,
            _input_image[0],
            NULL, 0, NULL
        );
    err = clEnqueueWriteImage
        (
            _info.command_queue,
            _input_green_image_id,
            CL_FALSE,
            img_index, img_region,
            0, 0,
            _input_image[1],
            NULL, 0, NULL
        );
    err = clEnqueueWriteImage
        (
            _info.command_queue,
            _input_blue_image_id,
            CL_FALSE,
            img_index, img_region,
            0, 0,
            _input_image[2],
            NULL, 0, NULL
        );

    // Wait for all commands to finish
    err = clEnqueueBarrierWithWaitList
        (
            _info.command_queue,
            0, NULL, NULL
        );
}

void grayscale_compute()
{
    cl_int err;
    const size_t img_origin[3] = {0, 0, 0};
    const size_t img_region[3] = {COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT, 1};
    cl_kernel kernel = clCreateKernel(_program, "img_grayscale", NULL);

    err = clSetKernelArg(kernel, 0, sizeof(cl_mem), (void *) &_input_red_image_id);
    err = clSetKernelArg(kernel, 1, sizeof(cl_mem), (void *) &_input_green_image_id);
    err = clSetKernelArg(kernel, 2, sizeof(cl_mem), (void *) &_input_blue_image_id);
    err = clSetKernelArg(kernel, 3, sizeof(cl_mem), (void *) &_output_image_id);

    // Dispatch the computing
    size_t global_item_size[2] = {COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT};
    size_t local_item_size[2] = {960, 1};

    clEnqueueAcquireGLObjects(_info.command_queue, 1, &(_output_image_id), 0, 0, NULL);
    err = clEnqueueNDRangeKernel
        (
            _info.command_queue,
            kernel,
            2,
            NULL,
            global_item_size,
            local_item_size,
            NULL, 0, NULL
        );
    clFinish(_info.command_queue);
    clEnqueueReleaseGLObjects(_info.command_queue, 1, &_output_image_id, 0, 0, NULL);
}

void grayscale_destroy()
{
    clReleaseMemObject(_input_red_image_id);
    clReleaseMemObject(_input_green_image_id);
    clReleaseMemObject(_input_blue_image_id);
    clReleaseMemObject(_output_image_id);
    clReleaseProgram(_program);
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_GL_FUNCTIONS_H
#define OPENINGS_GL_FUNCTIONS_H

/**
 * @brief Manual loading of the OpenGL 3.1 features, that are not included in the Android NDK
 * libGLESv3.so. Use it as a complement.
 */

#include <EGL/egl.h>
#include <GLES3/gl3.h>

#ifndef APIENTRY
#define APIENTRY
#endif
#ifndef APIENTRYP
#define APIENTRYP APIENTRY *
#endif

#ifndef GLAPIENTRY
#define GLAPIENTRY APIENTRY
#endif

#ifndef GL_ES_VERSION_3_1
#define GL_ES_VERSION_3_1 1
#endif

#define GL_COMPUTE_SHADER                                   0x91B9
#define GL_MAX_COMPUTE_UNIFORM_BLOCKS                       0x91BB
#define GL_MAX_COMPUTE_TEXTURE_IMAGE_UNITS                  0x91BC
#define GL_MAX_COMPUTE_IMAGE_UNIFORMS                       0x91BD
#define GL_MAX_COMPUTE_SHARED_MEMORY_SIZE                   0x8262
#define GL_MAX_COMPUTE_UNIFORM_COMPONENTS                   0x8263
#define GL_MAX_COMPUTE_ATOMIC_COUNTER_BUFFERS               0x8264
#define GL_MAX_COMPUTE_ATOMIC_COUNTERS                      0x8265
#define GL_MAX_COMBINED_COMPUTE_UNIFORM_COMPONENTS          0x8266
#define GL_MAX_COMPUTE_WORK_GROUP_INVOCATIONS               0x90EB
#define GL_MAX_COMPUTE_WORK_GROUP_COUNT                     0x91BE
#define GL_MAX_COMPUTE_WORK_GROUP_SIZE                      0x91BF
#define GL_COMPUTE_WORK_GROUP_SIZE                          0x8267
#define GL_DISPATCH_INDIRECT_BUFFER                         0x90EE
#define GL_DISPATCH_INDIRECT_BUFFER_BINDING                 0x90EF
#define GL_COMPUTE_SHADER_BIT                               0x00000020
#define GL_DRAW_INDIRECT_BUFFER                             0x8F3F
#define GL_DRAW_INDIRECT_BUFFER_BINDING                     0x8F43
#define GL_MAX_UNIFORM_LOCATIONS                            0x826E
#define GL_FRAMEBUFFER_DEFAULT_WIDTH                        0x9310
#define GL_FRAMEBUFFER_DEFAULT_HEIGHT                       0x9311
#define GL_FRAMEBUFFER_DEFAULT_SAMPLES                      0x9313
#define GL_FRAMEBUFFER_DEFAULT_FIXED_SAMPLE_LOCATIONS       0x9314
#define GL_MAX_FRAMEBUFFER_WIDTH                            0x9315
#define GL_MAX_FRAMEBUFFER_HEIGHT                           0x9316
#define GL_MAX_FRAMEBUFFER_SAMPLES                          0x9318
#define GL_UNIFORM                                          0x92E1
#define GL_UNIFORM_BLOCK                                    0x92E2
#define GL_PROGRAM_INPUT                                    0x92E3
#define GL_PROGRAM_OUTPUT                                   0x92E4
#define GL_BUFFER_VARIABLE                                  0x92E5
#define GL_SHADER_STORAGE_BLOCK                             0x92E6
#define GL_ATOMIC_COUNTER_BUFFER                            0x92C0
#define GL_TRANSFORM_FEEDBACK_VARYING                       0x92F4
#define GL_ACTIVE_RESOURCES                                 0x92F5
#define GL_MAX_NAME_LENGTH                                  0x92F6
#define GL_MAX_NUM_ACTIVE_VARIABLES                         0x92F7
#define GL_NAME_LENGTH                                      0x92F9
#define GL_TYPE                                             0x92FA
#define GL_ARRAY_SIZE                                       0x92FB
#define GL_OFFSET                                           0x92FC
#define GL_BLOCK_INDEX                                      0x92FD
#define GL_ARRAY_STRIDE                                     0x92FE
#define GL_MATRIX_STRIDE                                    0x92FF
#define GL_IS_ROW_MAJOR                                     0x9300
#define GL_ATOMIC_COUNTER_BUFFER_INDEX                      0x9301
#define GL_BUFFER_BINDING                                   0x9302
#define GL_BUFFER_DATA_SIZE                                 0x9303
#define GL_NUM_ACTIVE_VARIABLES                             0x9304
#define GL_ACTIVE_VARIABLES                                 0x9305
#define GL_REFERENCED_BY_VERTEX_SHADER                      0x9306
#define GL_REFERENCED_BY_FRAGMENT_SHADER                    0x930A
#define GL_REFERENCED_BY_COMPUTE_SHADER                     0x930B
#define GL_TOP_LEVEL_ARRAY_SIZE                             0x930C
#define GL_TOP_LEVEL_ARRAY_STRIDE                           0x930D
#define GL_LOCATION                                         0x930E
#define GL_VERTEX_SHADER_BIT                                0x00000001
#define GL_FRAGMENT_SHADER_BIT                              0x00000002
#define GL_ALL_SHADER_BITS                                  0xFFFFFFFF
#define GL_PROGRAM_SEPARABLE                                0x8258
#define GL_ACTIVE_PROGRAM                                   0x8259
#define GL_PROGRAM_PIPELINE_BINDING                         0x825A
#define GL_ATOMIC_COUNTER_BUFFER_BINDING                    0x92C1
#define GL_ATOMIC_COUNTER_BUFFER_START                      0x92C2
#define GL_ATOMIC_COUNTER_BUFFER_SIZE                       0x92C3
#define GL_MAX_VERTEX_ATOMIC_COUNTER_BUFFERS                0x92CC
#define GL_MAX_FRAGMENT_ATOMIC_COUNTER_BUFFERS              0x92D0
#define GL_MAX_COMBINED_ATOMIC_COUNTER_BUFFERS              0x92D1
#define GL_MAX_VERTEX_ATOMIC_COUNTERS                       0x92D2
#define GL_MAX_FRAGMENT_ATOMIC_COUNTERS                     0x92D6
#define GL_MAX_COMBINED_ATOMIC_COUNTERS                     0x92D7
#define GL_MAX_ATOMIC_COUNTER_BUFFER_SIZE                   0x92D8
#define GL_MAX_ATOMIC_COUNTER_BUFFER_BINDINGS               0x92DC
#define GL_ACTIVE_ATOMIC_COUNTER_BUFFERS                    0x92D9
#define GL_UNSIGNED_INT_ATOMIC_COUNTER                      0x92DB
#define GL_MAX_IMAGE_UNITS                                  0x8F38
#define GL_MAX_VERTEX_IMAGE_UNIFORMS                        0x90CA
#define GL_MAX_FRAGMENT_IMAGE_UNIFORMS                      0x90CE
#define GL_MAX_COMBINED_IMAGE_UNIFORMS                      0x90CF
#define GL_IMAGE_BINDING_NAME                               0x8F3A
#define GL_IMAGE_BINDING_LEVEL                              0x8F3B
#define GL_IMAGE_BINDING_LAYERED                            0x8F3C
#define GL_IMAGE_BINDING_LAYER                              0x8F3D
#define GL_IMAGE_BINDING_ACCESS                             0x8F3E
#define GL_IMAGE_BINDING_FORMAT                             0x906E
#define GL_VERTEX_ATTRIB_ARRAY_BARRIER_BIT                  0x00000001
#define GL_ELEMENT_ARRAY_BARRIER_BIT                        0x00000002
#define GL_UNIFORM_BARRIER_BIT                              0x00000004
#define GL_TEXTURE_FETCH_BARRIER_BIT                        0x00000008
#define GL_SHADER_IMAGE_ACCESS_BARRIER_BIT                  0x00000020
#define GL_COMMAND_BARRIER_BIT                              0x00000040
#define GL_PIXEL_BUFFER_BARRIER_BIT                         0x00000080
#define GL_TEXTURE_UPDATE_BARRIER_BIT                       0x00000100
#define GL_BUFFER_UPDATE_BARRIER_BIT                        0x00000200
#define GL_FRAMEBUFFER_BARRIER_BIT                          0x00000400
#define GL_TRANSFORM_FEEDBACK_BARRIER_BIT                   0x00000800
#define GL_ATOMIC_COUNTER_BARRIER_BIT                       0x00001000
#define GL_ALL_BARRIER_BITS                                 0xFFFFFFFF
#define GL_IMAGE_2D                                         0x904D
#define GL_IMAGE_3D                                         0x904E
#define GL_IMAGE_CUBE                                       0x9050
#define GL_IMAGE_2D_ARRAY                                   0x9053
#define GL_INT_IMAGE_2D                                     0x9058
#define GL_INT_IMAGE_3D                                     0x9059
#define GL_INT_IMAGE_CUBE                                   0x905B
#define GL_INT_IMAGE_2D_ARRAY                               0x905E
#define GL_UNSIGNED_INT_IMAGE_2D                            0x9063
#define GL_UNSIGNED_INT_IMAGE_3D                            0x9064
#define GL_UNSIGNED_INT_IMAGE_CUBE                          0x9066
#define GL_UNSIGNED_INT_IMAGE_2D_ARRAY                      0x9069
#define GL_IMAGE_FORMAT_COMPATIBILITY_TYPE                  0x90C7
#define GL_IMAGE_FORMAT_COMPATIBILITY_BY_SIZE               0x90C8
#define GL_IMAGE_FORMAT_COMPATIBILITY_BY_CLASS              0x90C9
#define GL_READ_ONLY                                        0x88B8
#define GL_WRITE_ONLY                                       0x88B9
#define GL_READ_WRITE                                       0x88BA
#define GL_SHADER_STORAGE_BUFFER                            0x90D2
#define GL_SHADER_STORAGE_BUFFER_BINDING                    0x90D3
#define GL_SHADER_STORAGE_BUFFER_START                      0x90D4
#define GL_SHADER_STORAGE_BUFFER_SIZE                       0x90D5
#define GL_MAX_VERTEX_SHADER_STORAGE_BLOCKS                 0x90D6
#define GL_MAX_FRAGMENT_SHADER_STORAGE_BLOCKS               0x90DA
#define GL_MAX_COMPUTE_SHADER_STORAGE_BLOCKS                0x90DB
#define GL_MAX_COMBINED_SHADER_STORAGE_BLOCKS               0x90DC
#define GL_MAX_SHADER_STORAGE_BUFFER_BINDINGS               0x90DD
#define GL_MAX_SHADER_STORAGE_BLOCK_SIZE                    0x90DE
#define GL_SHADER_STORAGE_BUFFER_OFFSET_ALIGNMENT           0x90DF
#define GL_SHADER_STORAGE_BARRIER_BIT                       0x00002000
#define GL_MAX_COMBINED_SHADER_OUTPUT_RESOURCES             0x8F39
#define GL_DEPTH_STENCIL_TEXTURE_MODE                       0x90EA
#define GL_STENCIL_INDEX                                    0x1901
#define GL_MIN_PROGRAM_TEXTURE_GATHER_OFFSET                0x8E5E
#define GL_MAX_PROGRAM_TEXTURE_GATHER_OFFSET                0x8E5F
#define GL_SAMPLE_POSITION                                  0x8E50
#define GL_SAMPLE_MASK                                      0x8E51
#define GL_SAMPLE_MASK_VALUE                                0x8E52
#define GL_TEXTURE_2D_MULTISAMPLE                           0x9100
#define GL_MAX_SAMPLE_MASK_WORDS                            0x8E59
#define GL_MAX_COLOR_TEXTURE_SAMPLES                        0x910E
#define GL_MAX_DEPTH_TEXTURE_SAMPLES                        0x910F
#define GL_MAX_INTEGER_SAMPLES                              0x9110
#define GL_TEXTURE_BINDING_2D_MULTISAMPLE                   0x9104
#define GL_TEXTURE_SAMPLES                                  0x9106
#define GL_TEXTURE_FIXED_SAMPLE_LOCATIONS                   0x9107
#define GL_TEXTURE_WIDTH                                    0x1000
#define GL_TEXTURE_HEIGHT                                   0x1001
#define GL_TEXTURE_DEPTH                                    0x8071
#define GL_TEXTURE_INTERNAL_FORMAT                          0x1003
#define GL_TEXTURE_RED_SIZE                                 0x805C
#define GL_TEXTURE_GREEN_SIZE                               0x805D
#define GL_TEXTURE_BLUE_SIZE                                0x805E
#define GL_TEXTURE_ALPHA_SIZE                               0x805F
#define GL_TEXTURE_DEPTH_SIZE                               0x884A
#define GL_TEXTURE_STENCIL_SIZE                             0x88F1
#define GL_TEXTURE_SHARED_SIZE                              0x8C3F
#define GL_TEXTURE_RED_TYPE                                 0x8C10
#define GL_TEXTURE_GREEN_TYPE                               0x8C11
#define GL_TEXTURE_BLUE_TYPE                                0x8C12
#define GL_TEXTURE_ALPHA_TYPE                               0x8C13
#define GL_TEXTURE_DEPTH_TYPE                               0x8C16
#define GL_TEXTURE_COMPRESSED                               0x86A1
#define GL_SAMPLER_2D_MULTISAMPLE                           0x9108
#define GL_INT_SAMPLER_2D_MULTISAMPLE                       0x9109
#define GL_UNSIGNED_INT_SAMPLER_2D_MULTISAMPLE              0x910A
#define GL_VERTEX_ATTRIB_BINDING                            0x82D4
#define GL_VERTEX_ATTRIB_RELATIVE_OFFSET                    0x82D5
#define GL_VERTEX_BINDING_DIVISOR                           0x82D6
#define GL_VERTEX_BINDING_OFFSET                            0x82D7
#define GL_VERTEX_BINDING_STRIDE                            0x82D8
#define GL_VERTEX_BINDING_BUFFER                            0x8F4F
#define GL_MAX_VERTEX_ATTRIB_RELATIVE_OFFSET                0x82D9
#define GL_MAX_VERTEX_ATTRIB_BINDINGS                       0x82DA
#define GL_MAX_VERTEX_ATTRIB_STRIDE                         0x82E5

// OpenGL functions names

#define glDispatchCompute               __glDispatchCompute
#define glDispatchComputeIndirect       __glDispatchComputeIndirect
#define glDrawArraysIndirect            __glDrawArraysIndirect
#define glDrawElementsIndirect          __glDrawElementsIndirect
#define glFramebufferParameteri         __glFramebufferParameteri
#define glGetFramebufferParameteri      __glGetFramebufferParameteri
#define glGetProgramInterface           __glGetProgramInterface
#define glGetProgramSourceIndex         __glGetProgramSourceIndex
#define glGetProgramResourceName        __glGetProgramResourceName
#define glGetProgramResourceiv          __glGetProgramResourceiv
#define glGetProgramResourceLocation    __glGetProgramResourceLocation
#define glUseProgramStages              __glUseProgramStages
#define glActiveShaderProgram           __glActiveShaderProgram
#define glCreateSHaderProgram           __glCreateSHaderProgram
#define glBindProgramPipeline           __glBindProgramPipeline
#define glDeleteProgramPipeline         __glDeleteProgramPipeline
#define glGenProgramPipeline            __glGenProgramPipeline
#define glIsProgramPipeline             __glIsProgramPipeline
#define glGetProgramPipelineiv          __glGetProgramPipelineiv
#define glProgramUniform1i              __glProgramUniform1i
#define glProgramUniform2i              __glProgramUniform2i
#define glProgramUniform3i              __glProgramUniform3i
#define glProgramUniform4i              __glProgramUniform4i
#define glProgramUniform1ui             __glProgramUniform1ui
#define glProgramUniform2ui             __glProgramUniform2ui
#define glProgramUniform3ui             __glProgramUniform3ui
#define glProgramUniform4ui             __glProgramUniform4ui
#define glProgramUniform1f              __glProgramUniform1f
#define glProgramUniform2f              __glProgramUniform2f
#define glProgramUniform3f              __glProgramUniform3f
#define glProgramUniform4f              __glProgramUniform4f
#define glProgramUniform1iv             __glProgramUniform1iv
#define glProgramUniform2iv             __glProgramUniform2iv
#define glProgramUniform3iv             __glProgramUniform3iv
#define glProgramUniform4iv             __glProgramUniform4iv
#define glProgramUniform1uiv            __glProgramUniform1uiv
#define glProgramUniform2uiv            __glProgramUniform2uiv
#define glProgramUniform3uiv            __glProgramUniform3uiv
#define glProgramUniform4uiv            __glProgramUniform4uiv
#define glProgramUniform1fv             __glProgramUniform1fv
#define glProgramUniform2fv             __glProgramUniform2fv
#define glProgramUniform3fv             __glProgramUniform3fv
#define glProgramUniform4fv             __glProgramUniform4fv
#define glProgramUniformMatrix2fv       __glProgramUniformMatrix2fv
#define glProgramUniformMatrix3fv       __glProgramUniformMatrix3fv
#define glProgramUniformMatrix4fv       __glProgramUniformMatrix4fv
#define glProgramUniformMatrix2x3fv     __glProgramUniformMatrix2x3fv
#define glProgramUniformMatrix3x2fv     __glProgramUniformMatrix3x2fv
#define glProgramUniformMatrix2x4fv     __glProgramUniformMatrix2x4fv
#define glProgramUniformMatrix4x2fv     __glProgramUniformMatrix4x2fv
#define glProgramUniformMatrix3x4fv     __glProgramUniformMatrix3x4fv
#define glProgramUniformMatrix4x3fv     __glProgramUniformMatrix4x3fv
#define glValidateProgramPipeline       __glValidateProgramPipeline
#define glGetProgramPipelineInfoLog     __glGetProgramPipelineInfoLog
#define glBindImageTexture              __glBindImageTexture
#define glGetBooleani_v                 __glGetBooleani_v
#define glMemoryBarrier                 __glMemoryBarrier
#define glMemoryBarrierByRegion         __glMemoryBarrierByRegion
#define glTexStorage2DMultiSample       __glTexStorage2DMultiSample
#define glGetMultisamplefv              __glGetMultisamplefv
#define glSampleMaski                   __glSampleMaski
#define glGetTexLevelParameteriv        __glGetTexLevelParameteriv
#define glGetTexLevelParameterfv        __glGetTexLevelParameterfv
#define glBindVertexBuffer              __glBindVertexBuffer
#define glVertexAttribFormat            __glVertexAttribFormat
#define glVertexAttribIFormatProc       __glVertexAttribIFormatProc
#define glVertexAttribBinding           __glVertexAttribBinding
#define glVertexBindingDivisor          __glVertexBindingDivisor

// -------------------------------------------------------------------------------------------------

__BEGIN_DECLS

typedef void (GL_APIENTRYP PFNGLDISPATCHCOMPUTEPROC)
    (
        GLuint num_groups_x,
        GLuint num_groups_y,
        GLuint num_groups_z
    );

typedef void (GL_APIENTRYP PFNGLDISPATCHCOMPUTEINDIRECTPROC)(GLintptr indirect);

typedef void (GL_APIENTRYP PFNGLDRAWARRAYSINDIRECTPROC)(GLenum mode, const void *indirect);

typedef void (GL_APIENTRYP PFNGLDRAWELEMENTSINDIRECTPROC)
    (
        GLenum mode,
        GLenum type,
        const void *indirect
    );

typedef void (GL_APIENTRYP PFNGLFRAMEBUFFERPARAMETERIPROC)
    (
        GLenum target,
        GLenum pname,
        GLint param
    );

typedef void (GL_APIENTRYP PFNGLGETFRAMEBUFFERPARAMETERIVPROC)
    (
        GLenum target,
        GLenum pname,
        GLint *params
    );

typedef void (GL_APIENTRYP PFNGLGETPROGRAMINTERFACEIVPROC)
    (
        GLuint program,
        GLenum programInterface,
        GLenum pname,
        GLint *params
    );

typedef GLuint (GL_APIENTRYP PFNGLGETPROGRAMRESOURCEINDEXPROC)
    (
        GLuint program,
        GLenum programInterface,
        const GLchar *name
    );

typedef void (GL_APIENTRYP PFNGLGETPROGRAMRESOURCENAMEPROC)
    (
        GLuint program,
        GLenum programInterface,
        GLuint index,
        GLsizei bufSize,
        GLsizei *length,
        GLchar *name
    );

typedef void (GL_APIENTRYP PFNGLGETPROGRAMRESOURCEIVPROC)
    (
        GLuint program,
        GLenum programInterface,
        GLuint index,
        GLsizei propCount,
        const GLenum *props,
        GLsizei bufSize,
        GLsizei *length,
        GLint *params);

typedef GLint (GL_APIENTRYP PFNGLGETPROGRAMRESOURCELOCATIONPROC)
    (
        GLuint program,
        GLenum programInterface,
        const GLchar *name
    );

typedef void (GL_APIENTRYP PFNGLUSEPROGRAMSTAGESPROC)
    (
        GLuint pipeline,
        GLbitfield stages,
        GLuint program
    );

typedef void (GL_APIENTRYP PFNGLACTIVESHADERPROGRAMPROC)(GLuint pipeline, GLuint program);

typedef GLuint (GL_APIENTRYP PFNGLCREATESHADERPROGRAMVPROC)
    (
        GLenum type,
        GLsizei count,
        const GLchar *const *strings
    );

typedef void (GL_APIENTRYP PFNGLBINDPROGRAMPIPELINEPROC)(GLuint pipeline);

typedef void (GL_APIENTRYP PFNGLDELETEPROGRAMPIPELINESPROC)(GLsizei n, const GLuint *pipelines);

typedef void (GL_APIENTRYP PFNGLGENPROGRAMPIPELINESPROC)(GLsizei n, GLuint *pipelines);

typedef GLboolean (GL_APIENTRYP PFNGLISPROGRAMPIPELINEPROC)(GLuint pipeline);

typedef void (GL_APIENTRYP PFNGLGETPROGRAMPIPELINEIVPROC)
    (
        GLuint pipeline,
        GLenum pname,
        GLint *params
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM1IPROC)
    (
        GLuint program,
        GLint location,
        GLint v0
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM2IPROC)
    (
        GLuint program,
        GLint location,
        GLint v0,
        GLint v1
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM3IPROC)
    (
        GLuint program,
        GLint location,
        GLint v0,
        GLint v1,
        GLint v2
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM4IPROC)
    (
        GLuint program,
        GLint location,
        GLint v0,
        GLint v1,
        GLint v2,
        GLint v3
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM1UIPROC)
    (
        GLuint program,
        GLint location,
        GLuint v0
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM2UIPROC)
    (
        GLuint program,
        GLint location,
        GLuint v0,
        GLuint v1
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM3UIPROC)
    (
        GLuint program,
        GLint location,
        GLuint v0,
        GLuint v1,
        GLuint v2
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM4UIPROC)
    (
        GLuint program,
        GLint location,
        GLuint v0,
        GLuint v1,
        GLuint v2,
        GLuint v3
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM1FPROC)
    (
        GLuint program,
        GLint location,
        GLfloat v0
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM2FPROC)
    (
        GLuint program,
        GLint location,
        GLfloat v0,
        GLfloat v1
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM3FPROC)
    (
        GLuint program,
        GLint location,
        GLfloat v0,
        GLfloat v1,
        GLfloat v2
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM4FPROC)
    (
        GLuint program,
        GLint location,
        GLfloat v0,
        GLfloat v1,
        GLfloat v2,
        GLfloat v3
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM1IVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM2IVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM3IVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM4IVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM1UIVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLuint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM2UIVPROC)
    (GLuint program,
     GLint location,
     GLsizei count,
     const GLuint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM3UIVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLuint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM4UIVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLuint *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM1FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM2FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM3FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORM4FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX2FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX3FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX4FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX2X3FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX3X2FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX2X4FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX4X2FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX3X4FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLPROGRAMUNIFORMMATRIX4X3FVPROC)
    (
        GLuint program,
        GLint location,
        GLsizei count,
        GLboolean transpose,
        const GLfloat *value
    );

typedef void (GL_APIENTRYP PFNGLVALIDATEPROGRAMPIPELINEPROC)(GLuint pipeline);

typedef void (GL_APIENTRYP PFNGLGETPROGRAMPIPELINEINFOLOGPROC)
    (
        GLuint pipeline,
        GLsizei bufSize,
        GLsizei *length,
        GLchar *infoLog
    );

typedef void (GL_APIENTRYP PFNGLBINDIMAGETEXTUREPROC)
    (
        GLuint unit,
        GLuint texture,
        GLint level,
        GLboolean layered,
        GLint layer, GLenum access,
        GLenum format
    );

typedef void (GL_APIENTRYP PFNGLGETBOOLEANI_VPROC)(GLenum target, GLuint index, GLboolean *data);

typedef void (GL_APIENTRYP PFNGLMEMORYBARRIERPROC)(GLbitfield barriers);

typedef void (GL_APIENTRYP PFNGLMEMORYBARRIERBYREGIONPROC)(GLbitfield barriers);

typedef void (GL_APIENTRYP PFNGLTEXSTORAGE2DMULTISAMPLEPROC)
    (
        GLenum target,
        GLsizei samples,
        GLenum internalformat,
        GLsizei width,
        GLsizei height,
        GLboolean fixedsamplelocations
    );

typedef void (GL_APIENTRYP PFNGLGETMULTISAMPLEFVPROC)(GLenum pname, GLuint index, GLfloat *val);

typedef void (GL_APIENTRYP PFNGLSAMPLEMASKIPROC)(GLuint maskNumber, GLbitfield mask);

typedef void (GL_APIENTRYP PFNGLGETTEXLEVELPARAMETERIVPROC)
    (
        GLenum target,
        GLint level,
        GLenum pname,
        GLint *params
    );

typedef void (GL_APIENTRYP PFNGLGETTEXLEVELPARAMETERFVPROC)
    (
        GLenum target,
        GLint level,
        GLenum pname,
        GLfloat *params
    );

typedef void (GL_APIENTRYP PFNGLBINDVERTEXBUFFERPROC)
    (
        GLuint bindingindex,
        GLuint buffer,
        GLintptr offset,
        GLsizei stride
    );

typedef void (GL_APIENTRYP PFNGLVERTEXATTRIBFORMATPROC)
    (
        GLuint attribindex,
        GLint size,
        GLenum type,
        GLboolean normalized,
        GLuint relativeoffset
    );

typedef void (GL_APIENTRYP PFNGLVERTEXATTRIBIFORMATPROC)
    (
        GLuint attribindex,
        GLint size,
        GLenum type,
        GLuint relativeoffset
    );

typedef void (GL_APIENTRYP PFNGLVERTEXATTRIBBINDINGPROC)(GLuint attribindex, GLuint bindingindex);

typedef void (GL_APIENTRYP PFNGLVERTEXBINDINGDIVISORPROC)(GLuint bindingindex, GLuint divisor);

// -------------------------------------------------------------------------------------------------

/**
 * Loads the OpenGLES 3.1 function pointers
 */
extern void gl_functions_load();

extern PFNGLDISPATCHCOMPUTEPROC __glDispatchCompute;
extern PFNGLDISPATCHCOMPUTEINDIRECTPROC __glDispatchComputeIndirect;
extern PFNGLDRAWARRAYSINDIRECTPROC __glDrawArraysIndirect;
extern PFNGLDRAWELEMENTSINDIRECTPROC __glDrawElementsIndirect;
extern PFNGLFRAMEBUFFERPARAMETERIPROC __glFramebufferParameteri;
extern PFNGLGETFRAMEBUFFERPARAMETERIVPROC __glGetFramebufferParameteri;
extern PFNGLGETPROGRAMINTERFACEIVPROC __glGetProgramInterface;
extern PFNGLGETPROGRAMRESOURCEINDEXPROC __glGetProgramSourceIndex;
extern PFNGLGETPROGRAMRESOURCENAMEPROC __glGetProgramResourceName;
extern PFNGLGETPROGRAMRESOURCEIVPROC __glGetProgramResourceiv;
extern PFNGLGETPROGRAMRESOURCELOCATIONPROC __glGetProgramResourceLocation;
extern PFNGLUSEPROGRAMSTAGESPROC __glUseProgramStages;
extern PFNGLACTIVESHADERPROGRAMPROC __glActiveShaderProgram;
extern PFNGLCREATESHADERPROGRAMVPROC __glCreateShaderProgram;
extern PFNGLBINDPROGRAMPIPELINEPROC __glBindProgramPipeline;
extern PFNGLDELETEPROGRAMPIPELINESPROC __glDeleteProgramPipeline;
extern PFNGLGENPROGRAMPIPELINESPROC __glGenProgramPipeline;
extern PFNGLISPROGRAMPIPELINEPROC __glIsProgramPipeline;
extern PFNGLGETPROGRAMPIPELINEIVPROC __glGetProgramPipelineiv;
extern PFNGLPROGRAMUNIFORM1IPROC __glProgramUniform1i;
extern PFNGLPROGRAMUNIFORM2IPROC __glProgramUniform2i;
extern PFNGLPROGRAMUNIFORM3IPROC __glProgramUniform3i;
extern PFNGLPROGRAMUNIFORM4IPROC __glProgramUniform4i;
extern PFNGLPROGRAMUNIFORM1UIPROC __glProgramUniform1ui;
extern PFNGLPROGRAMUNIFORM2UIPROC __glProgramUniform2ui;
extern PFNGLPROGRAMUNIFORM3UIPROC __glProgramUniform3ui;
extern PFNGLPROGRAMUNIFORM4UIPROC __glProgramUniform4ui;
extern PFNGLPROGRAMUNIFORM1FPROC __glProgramUniform1f;
extern PFNGLPROGRAMUNIFORM2FPROC __glProgramUniform2f;
extern PFNGLPROGRAMUNIFORM3FPROC __glProgramUniform3f;
extern PFNGLPROGRAMUNIFORM4FPROC __glProgramUniform4f;
extern PFNGLPROGRAMUNIFORM1IVPROC __glProgramUniform1iv;
extern PFNGLPROGRAMUNIFORM2IVPROC __glProgramUniform2iv;
extern PFNGLPROGRAMUNIFORM3IVPROC __glProgramUniform3iv;
extern PFNGLPROGRAMUNIFORM4IVPROC __glProgramUniform4iv;
extern PFNGLPROGRAMUNIFORM1UIVPROC __glProgramUniform1uiv;
extern PFNGLPROGRAMUNIFORM2UIVPROC __glProgramUniform2uiv;
extern PFNGLPROGRAMUNIFORM3UIVPROC __glProgramUniform3uiv;
extern PFNGLPROGRAMUNIFORM4UIVPROC __glProgramUniform4uiv;
extern PFNGLPROGRAMUNIFORM1FVPROC __glProgramUniform1fv;
extern PFNGLPROGRAMUNIFORM2FVPROC __glProgramUniform2fv;
extern PFNGLPROGRAMUNIFORM3FVPROC __glProgramUniform3fv;
extern PFNGLPROGRAMUNIFORM4FVPROC __glProgramUniform4fv;
extern PFNGLPROGRAMUNIFORMMATRIX2FVPROC __glProgramUniformMatrix2fv;
extern PFNGLPROGRAMUNIFORMMATRIX3FVPROC __glProgramUniformMatrix3fv;
extern PFNGLPROGRAMUNIFORMMATRIX4FVPROC __glProgramUniformMatrix4fv;
extern PFNGLPROGRAMUNIFORMMATRIX2X3FVPROC __glProgramUniformMatrix2x3fv;
extern PFNGLPROGRAMUNIFORMMATRIX3X2FVPROC __glProgramUniformMatrix3x2fv;
extern PFNGLPROGRAMUNIFORMMATRIX2X4FVPROC __glProgramUniformMatrix2x4fv;
extern PFNGLPROGRAMUNIFORMMATRIX4X2FVPROC __glProgramUniformMatrix4x2fv;
extern PFNGLPROGRAMUNIFORMMATRIX3X4FVPROC __glProgramUniformMatrix3x4fv;
extern PFNGLPROGRAMUNIFORMMATRIX4X3FVPROC __glProgramUniformMatrix4x3fv;
extern PFNGLVALIDATEPROGRAMPIPELINEPROC __glValidateProgramPipeline;
extern PFNGLGETPROGRAMPIPELINEINFOLOGPROC __glGetProgramPipelineInfoLog;
extern PFNGLBINDIMAGETEXTUREPROC __glBindImageTexture;
extern PFNGLGETBOOLEANI_VPROC __glGetBooleani_v;
extern PFNGLMEMORYBARRIERPROC __glMemoryBarrier;
extern PFNGLMEMORYBARRIERBYREGIONPROC __glMemoryBarrierByRegion;
extern PFNGLTEXSTORAGE2DMULTISAMPLEPROC __glTexStorage2DMultiSample;
extern PFNGLGETMULTISAMPLEFVPROC __glGetMultisamplefv;
extern PFNGLSAMPLEMASKIPROC __glSampleMaski;
extern PFNGLGETTEXLEVELPARAMETERIVPROC __glGetTexLevelParameteriv;
extern PFNGLGETTEXLEVELPARAMETERFVPROC __glGetTexLevelParameterfv;
extern PFNGLBINDVERTEXBUFFERPROC __glBindVertexBuffer;
extern PFNGLVERTEXATTRIBFORMATPROC __glVertexAttribFormat;
extern PFNGLVERTEXATTRIBIFORMATPROC __glVertexAttribIFormatProc;
extern PFNGLVERTEXATTRIBBINDINGPROC __glVertexAttribBinding;
extern PFNGLVERTEXBINDINGDIVISORPROC __glVertexBindingDivisor;

__END_DECLS

#endif // OPENINGS_GL_31_FUNCTIONS_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "gl_thread.h"

static struct compute_data _compute_data;
static struct compute_data *_compute_data_ptr;

static struct display_data _display_data;
static struct display_data *_display_data_ptr;

static pthread_t _gl_thread;
static uint8_t _status;
static uint8_t _stopped;

// -----------------------------------------------------------------------------------------

static void *__gl_thread_main_loop(void *);

static inline uint8_t __gl_thread__read_status()
{
    uint8_t res;
#pragma omp atomic capture
    {
        res = _status;
        _status = 0u;
    };

    return res;
}

// -----------------------------------------------------------------------------------------

void gl_thread_set_flag(uint8_t flag)
{
#pragma omp atomic update
    _status |= flag;
}

void gl_thread_start(struct ANativeWindow *window)
{
    _compute_data_ptr = &_compute_data;
    _display_data_ptr = &_display_data;

#pragma omp atomic write
    _stopped = 0u;

    gl_context_create(window);

    if (pthread_create(&_gl_thread, NULL, &__gl_thread_main_loop, NULL))
    {
        LOGE("Problem initializing the GL thread : %s", strerror(errno));
    }
}

void gl_thread_stop()
{
#pragma omp atomic write
    _stopped = 1u;

    while (!(pthread_join(_gl_thread, NULL)))
    {
        LOGD("Stopping thread");
    }

    gl_context_destroy();
}

// -----------------------------------------------------------------------------------------

static void *__gl_thread_main_loop(void *context)
{
    uint8_t status;
    uint8_t stopped;

    gl_context_bind();
    gl_context_init_frame(_display_data_ptr);

#if(USE_OPENCL == 0)
    gl_context_init_computing(_compute_data_ptr);
#else
    cl_context_init_computing(_compute_data_ptr);
#endif

    do
    {
        // Perform computings
        status = __gl_thread__read_status();
        model_get_compute_data(&_compute_data_ptr);

#if (USE_OPENCL == 0)
        if (status & GL_THREAD_COMPUTE_UPDATE_REQUEST)
            gl_context_update_computing(_compute_data_ptr);

        if (status & GL_THREAD_COMPUTE_PERFORM_REQUEST)
            gl_context_perform_computing(_compute_data_ptr);
#else
        if (status & GL_THREAD_COMPUTE_UPDATE_REQUEST)
            cl_context_update_computing(_compute_data_ptr);

        if (status & GL_THREAD_COMPUTE_PERFORM_REQUEST)
            cl_context_perform_computing(_compute_data_ptr);
#endif

        // Perform drawings
        model_get_display_data(&_display_data_ptr);

        if (status & GL_THREAD_RENDER_CLEAR_REQUEST)
            gl_context_clear_frame(_display_data_ptr);

        if (status & GL_THREAD_RENDER_RESIZE_REQUEST)
            gl_context_resize_frame(_display_data_ptr);

        if (status & GL_THREAD_RENDER_UPDATE_REQUEST)
            gl_context_update_frame(_display_data_ptr);

        gl_context_draw_frame(_display_data_ptr);

        usleep(GL_THREAD_SLEEP_TIME);

#pragma omp atomic read
        stopped = _stopped;

    } while (!stopped);

#if (USE_OPENCL == 0)
    gl_context_destroy_computing(_compute_data_ptr);
#else
    cl_context_destroy_computing(_compute_data_ptr);
#endif

    gl_context_destroy_frame(_display_data_ptr);
    gl_context_unbind();

    return NULL;
}<file_sep>/*
 * Created by <NAME> on 4/27/18 8:38 AM.
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 * along with this program.  If not, see <http://www.gnu.org/licenses/>.
 */

package limsi.fr.openings;

import android.app.Activity;
import android.hardware.Camera;
import android.os.Bundle;
import android.widget.FrameLayout;

/**
 * A basic activity that creates an UI with a NativeSurfaceView object and wraps all its
 * lifecycle ti native methods.
 */
public class MainActivity extends Activity {

    private Camera _camera;
    private CameraSurfaceView _cameraSurfaceView;
    private FrameLayout _cameraLayout;

    @Override
    protected void onCreate(Bundle savedInstanceState) {
        super.onCreate(savedInstanceState);
        setContentView(R.layout.activity_main);

        _camera = Camera.open();
        _cameraSurfaceView = new CameraSurfaceView(this, _camera);
        _cameraLayout = findViewById(R.id.camera_frame_layout);
        _cameraLayout.addView(_cameraSurfaceView);

        nativeOnCreate();
    }

    @Override
    protected void onStart() {
        super.onStart();
        nativeOnStart();
    }

    @Override
    protected void onResume() {
        super.onResume();
        nativeOnResume();
    }

    @Override
    protected void onPause() {
        super.onPause();
        nativeOnPause();
    }

    @Override
    protected void onStop() {
        super.onStop();
        nativeOnStop();
    }

    @Override
    protected void onDestroy() {
        super.onDestroy();
        nativeOnDestroy();
    }

    /*
     * Native methods for all the activity's lifecycle
     */
    private static native void nativeOnCreate();

    private static native void nativeOnStart();

    private static native void nativeOnResume();

    private static native void nativeOnPause();

    private static native void nativeOnStop();

    private static native void nativeOnDestroy();
}
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "application.h"

void application_create()
{
    model_init();
}

void application_init_window(struct ANativeWindow *window)
{
    gl_thread_start(window);
}

void application_term_window()
{
    gl_thread_stop();
}

void application_destroy()
{
    model_destroy();
}

void application_pause()
{

}

void application_resume()
{

}

void application_start()
{

}

void application_stop()
{

}

void application_redraw_needed()
{

}

void application_window_resized(int32_t width, int32_t height)
{
    gl_thread_set_flag(GL_THREAD_RENDER_RESIZE_REQUEST);
}
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include <stdio.h>
#include <stdlib.h>
#include <string.h>

#define LINE_MAX_LENGTH 512

int main(int argc, char **argv)
{
    char line[512];
    FILE *file;
    size_t nb_read;

    printf("/*\n"
               " * Created by <NAME>\n"
               " *\n"
               " * This program is free software: you can redistribute it and/or modify\n"
               " * it under the terms of the GNU General Public License as published by\n"
               " * the Free Software Foundation, either version 3 of the License, or\n"
               " * (at your option) any later version.\n"
               " *\n"
               " * This program is distributed in the hope that it will be useful,\n"
               " * but WITHOUT ANY WARRANTY; without even the implied warranty of\n"
               " * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the\n"
               " * GNU General Public License for more details.\n"
               " *\n"
               " * You should have received a copy of the GNU General Public License\n"
               " */\n\n");
    printf("#ifndef __KERNEL_SOURCES_H\n"
               "#define __KERNEL_SOURCES_H\n\n");

    for (int i = 1; i < argc; i++)
    {
        char filename[LINE_MAX_LENGTH];
        char *tmp;
        sprintf(filename, "%s.cl", argv[i]);
        file = fopen(filename, "r");
        printf("static const char *%s_source = \n", argv[i]);
        printf("\t{\n");
        while (fgets(&line, LINE_MAX_LENGTH, file))
        {
            tmp = strstr(line, "\n");
            if (tmp != NULL)
                strncpy(tmp, "\0", 2);
            printf("\t\t\"%s\\n\"\n", line);
        }
        printf("\t};\n\n");
    }
    printf("#endif // __KERNEL_SOURCES_H\n");
    return 0;
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "conv3x3_f.h"

static GLuint _conv3x3_f_compute_program_id;

static GLuint _conv3x3_f_red_texture_id;
static GLuint _conv3x3_f_green_texture_id;
static GLuint _conv3x3_f_blue_texture_id;

static GLuint _conv3x3_f_red_kernel_ubo_id;
static GLuint _conv3x3_f_green_kernel_ubo_id;
static GLuint _conv3x3_f_blue_kernel_ubo_id;

void conv3x3_f_init()
{
    // Image textures allocation
    glGenTextures(1, &_conv3x3_f_red_texture_id);
    glGenTextures(1, &_conv3x3_f_green_texture_id);
    glGenTextures(1, &_conv3x3_f_blue_texture_id);

    glBindTexture(GL_TEXTURE_2D, _conv3x3_f_red_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    glBindTexture(GL_TEXTURE_2D, _conv3x3_f_green_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    glBindTexture(GL_TEXTURE_2D, _conv3x3_f_blue_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    glBindTexture(GL_TEXTURE_2D, 0);

    // Kernels buffer allocation
    glGenBuffers(1, &_conv3x3_f_red_kernel_ubo_id);
    glGenBuffers(1, &_conv3x3_f_green_kernel_ubo_id);
    glGenBuffers(1, &_conv3x3_f_blue_kernel_ubo_id);

    glBindBuffer(GL_UNIFORM_BUFFER, _conv3x3_f_red_kernel_ubo_id);
    glBufferData
        (
            GL_UNIFORM_BUFFER,
            (GLsizeiptr) (3 * 3 * sizeof(GLfloat)),
            NULL,
            GL_DYNAMIC_DRAW
        );
    glBindBuffer(GL_UNIFORM_BUFFER, _conv3x3_f_green_kernel_ubo_id);
    glBufferData
        (
            GL_UNIFORM_BUFFER,
            (GLsizeiptr) (3 * 3 * sizeof(GLfloat)),
            NULL,
            GL_DYNAMIC_DRAW
        );
    glBindBuffer(GL_UNIFORM_BUFFER, _conv3x3_f_blue_kernel_ubo_id);
    glBufferData
        (
            GL_UNIFORM_BUFFER,
            (GLsizeiptr) (3 * 3 * sizeof(GLfloat)),
            NULL,
            GL_DYNAMIC_DRAW
        );
    glBindBuffer(GL_UNIFORM_BUFFER, 0);

    // Compute program creation
    _conv3x3_f_compute_program_id = shader_create_compute_program(CONV3X3_source);

    glUseProgram(_conv3x3_f_compute_program_id);
    glBindBufferBase(GL_UNIFORM_BUFFER, 0, _conv3x3_f_red_kernel_ubo_id);
    glBindBufferBase(GL_UNIFORM_BUFFER, 1, _conv3x3_f_green_kernel_ubo_id);
    glBindBufferBase(GL_UNIFORM_BUFFER, 2, _conv3x3_f_blue_kernel_ubo_id);
    glBindImageTexture(3, _conv3x3_f_red_texture_id, 0, GL_FALSE, 0, GL_READ_ONLY, GL_R32F);
    glBindImageTexture(4, _conv3x3_f_green_texture_id, 0, GL_FALSE, 0, GL_READ_ONLY, GL_R32F);
    glBindImageTexture(5, _conv3x3_f_blue_texture_id, 0, GL_FALSE, 0, GL_READ_ONLY, GL_R32F);
    glUseProgram(0);
}

void conv3x3_f_set_kernels(const float *kernels)
{
    glBindBuffer(GL_UNIFORM_BUFFER, _conv3x3_f_red_kernel_ubo_id);
    glBufferSubData(GL_UNIFORM_BUFFER, 0, (GLsizeiptr) (3 * 3 * sizeof(GLfloat)), kernels);
    glBindBuffer(GL_UNIFORM_BUFFER, _conv3x3_f_green_kernel_ubo_id);
    glBufferSubData(GL_UNIFORM_BUFFER, 0, (GLsizeiptr) (3 * 3 * sizeof(GLfloat)), kernels + 9);
    glBindBuffer(GL_UNIFORM_BUFFER, _conv3x3_f_blue_kernel_ubo_id);
    glBufferSubData(GL_UNIFORM_BUFFER, 0, (GLsizeiptr) (3 * 3 * sizeof(GLfloat)), kernels + 18);
    glBindBuffer(GL_UNIFORM_BUFFER, 0);
}

void conv3x3_f_attach_rendering_texture(const GLuint tex)
{
    glUseProgram(_conv3x3_f_compute_program_id);
    glBindImageTexture(6, tex, 0, GL_FALSE, 0, GL_WRITE_ONLY, GL_RGBA32F);
    glUseProgram(0);
}

void conv3x3_f_update(const float *red, const float *green, const float *blue)
{
    glBindTexture(GL_TEXTURE_2D, _conv3x3_f_red_texture_id);
    glTexSubImage2D
        (
            GL_TEXTURE_2D,
            0, 0, 0,
            COLOR_CAMERA_WIDTH,
            COLOR_CAMERA_HEIGHT,
            GL_RED,
            GL_FLOAT,
            red
        );

    glBindTexture(GL_TEXTURE_2D, _conv3x3_f_green_texture_id);
    glTexSubImage2D
        (
            GL_TEXTURE_2D,
            0, 0, 0,
            COLOR_CAMERA_WIDTH,
            COLOR_CAMERA_HEIGHT,
            GL_RED,
            GL_FLOAT,
            green
        );

    glBindTexture(GL_TEXTURE_2D, _conv3x3_f_blue_texture_id);
    glTexSubImage2D
        (
            GL_TEXTURE_2D,
            0, 0, 0,
            COLOR_CAMERA_WIDTH,
            COLOR_CAMERA_HEIGHT,
            GL_RED,
            GL_FLOAT,
            blue
        );
    glMemoryBarrier(GL_TEXTURE_UPDATE_BARRIER_BIT);
}

void conv3x3_f_perform_compute()
{
    glUseProgram(_conv3x3_f_compute_program_id);
    glDispatchCompute(COLOR_CAMERA_WIDTH, 2, 1);
    glMemoryBarrier(GL_SHADER_IMAGE_ACCESS_BARRIER_BIT);
    glUseProgram(0);
}

void conv3x3_f_destroy()
{
    glDeleteBuffers(1, &_conv3x3_f_red_kernel_ubo_id);
    glDeleteBuffers(1, &_conv3x3_f_green_kernel_ubo_id);
    glDeleteBuffers(1, &_conv3x3_f_blue_kernel_ubo_id);

    glDeleteTextures(1, &_conv3x3_f_red_texture_id);
    glDeleteTextures(1, &_conv3x3_f_green_texture_id);
    glDeleteTextures(1, &_conv3x3_f_blue_texture_id);

    glDeleteProgram(_conv3x3_f_compute_program_id);
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "kernel.h"

cl_program kernel_create_program
    (
        cl_context context,
        cl_device_id *device,
        const char *source
    )
{
    cl_program res = NULL;
    cl_int err;

    cl_build_status status;
    char *info_log;
    size_t log_len;

    res = clCreateProgramWithSource
        (
            context,
            1,
            (const char **) &source,
            NULL,
            &err
        );

    checkCLError("Creating cl program", err);
    if (err != CL_SUCCESS)
    {
        return NULL;
    }

    err = clBuildProgram(res, 1, device, NULL, NULL, NULL);
    err = clGetProgramBuildInfo
        (
            res,
            *device,
            CL_PROGRAM_BUILD_STATUS,
            sizeof(cl_program_build_info),
            &status,
            &log_len
        );

    if (status != CL_SUCCESS)
    {
        LOGE("OpenCL program building failed");
        err = clGetProgramBuildInfo
            (
                res, *device,
                CL_PROGRAM_BUILD_LOG,
                0,
                NULL,
                &log_len
            );

        info_log = (char *) malloc(log_len);
        clGetProgramBuildInfo
            (
                res, *device,
                CL_PROGRAM_BUILD_LOG,
                log_len,
                info_log,
                NULL
            );

        LOGE("%s", info_log);

        free(info_log);

        clReleaseProgram(res);
        res = NULL;
    }

    return res;
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_APPLICATION_H
#define OPENINGS_APPLICATION_H

/**
 * @brief Contains all the functions that handle the Android activity lifecycle.
 */

#include <stdlib.h>
#include <string.h>
#include <stdint.h>
#include <omp.h>

#include "global/global.h"
#include "gl/gl_thread.h"
#include "rendering/rendering.h"
#include "gl_computing/gl_computing.h"

__BEGIN_DECLS

extern void application_create();

extern void application_init_window(struct ANativeWindow *window);

extern void application_term_window();

extern void application_destroy();

extern void application_pause();

extern void application_resume();

extern void application_start();

extern void application_stop();

extern void application_redraw_needed();

extern void application_window_resized(int32_t width, int32_t height);

__END_DECLS

#endif // OPENINGS_APPLICATION_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include <stdio.h>
#include <stdlib.h>
#include <string.h>
#include <stdint.h>
#include <errno.h>

#include <jni.h>
#include <android/native_window.h>
#include <android/native_window_jni.h>

#include "global.h"
#include "application.h"

/**
 * This file contains the definition of all the native functions declared in the Java files.
 * These wrappers dispatch the Java calls into the associated native calls.
 */

__BEGIN_DECLS

static struct compute_data _compute_data;
static struct compute_data *_compute_data_ptr = &_compute_data;

static void conversions_convert_YUV_to_RGB
    (
        float *img,
        const uint8_t *yuv,
        const size_t width,
        const size_t height
    );

JNIEXPORT void JNICALL
Java_limsi_fr_openings_MainActivity_nativeOnCreate(JNIEnv *env, jobject obj)
{
    application_create();
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_MainActivity_nativeOnStart(JNIEnv *env, jobject obj)
{
    application_start();
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_MainActivity_nativeOnResume(JNIEnv *env, jobject obj)
{
    application_resume();
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_MainActivity_nativeOnPause(JNIEnv *env, jobject obj)
{
    application_pause();
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_MainActivity_nativeOnStop(JNIEnv *env, jobject obj)
{
    application_stop();
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_MainActivity_nativeOnDestroy(JNIEnv *env, jobject obj)
{
    application_destroy();
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_NativeSurfaceView_nativeCreateSurface
    (
        JNIEnv *env, jobject obj,
        jobject surface
    )
{
    struct ANativeWindow *window = ANativeWindow_fromSurface(env, surface);
    application_init_window(window);
    ANativeWindow_release(window);
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_NativeSurfaceView_nativeResizeSurface
    (
        JNIEnv *env, jobject obj,
        jobject surface,
        jint format,
        jint width,
        jint height
    )
{
    application_window_resized((int32_t) width, (int32_t) height);
}

JNIEXPORT void JNICALL
Java_limsi_fr_openings_NativeSurfaceView_nativeDrawSurface
    (
        JNIEnv *env, jobject obj,
        jobject surface
    )
{
    application_redraw_needed();
}


JNIEXPORT void JNICALL
Java_limsi_fr_openings_NativeSurfaceView_nativeDestroySurface
    (
        JNIEnv *env, jobject obj,
        jobject surface
    )
{
    application_term_window();
}

static uint8_t _img[COLOR_CAMERA_IMAGE_SIZE];

JNIEXPORT void JNICALL
Java_limsi_fr_openings_CameraSurfaceView_nativeSetImage
    (
        JNIEnv *env,
        jobject obj,
        jbyteArray img,
        jint width,
        jint height
    )
{
    jbyte *ptr = (*env)->GetByteArrayElements(env, img, NULL);
    jsize length = (*env)->GetArrayLength(env, img);

    conversions_convert_YUV_to_RGB(_compute_data_ptr->input_image, (uint8_t *) ptr,
                                   COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    (*env)->ReleaseByteArrayElements(env, img, ptr, 0);

    model_set_compute_data(&_compute_data_ptr);
    gl_thread_set_flag(GL_THREAD_COMPUTE_UPDATE_REQUEST);
    gl_thread_set_flag(GL_THREAD_COMPUTE_PERFORM_REQUEST);
    gl_thread_set_flag(GL_THREAD_RENDER_UPDATE_REQUEST);
}

static void conversions_convert_YUV_to_RGB
    (
        float *img,
        const uint8_t *yuv,
        const size_t width,
        const size_t height
    )
{
    const size_t frameSize = width * height;
    size_t a = 0;
    int y, u, v, r, g, b;

    size_t i, j;
    for (i = 0; i < height; i++)
    {
        for (j = 0; j < width; j++)
        {
            y = (0xff & ((int32_t) yuv[i * width + j]));
            v = (0xff & ((int32_t) yuv[frameSize + (i >> 1) * width + (j & ~1) + 0]));
            u = (0xff & ((int32_t) yuv[frameSize + (i >> 1) * width + (j & ~1) + 1]));
            y = y < 16 ? 16 : y;

            r = y + (int32_t) (1.370705f * (v - 128.));
            g = y - (int32_t) (0.698001f * (v - 128.)) - (int32_t) (0.337633 * (u - 128.));
            b = y + (int32_t) (1.732446f * (u - 128.));

            r = r < 0 ? 0 : (r > 255 ? 255 : r);
            g = g < 0 ? 0 : (g > 255 ? 255 : g);
            b = b < 0 ? 0 : (b > 255 ? 255 : b);
            a = 3 * (i * width + j);

            img[a] = ((float) r / 255.0f);
            img[a + 1] = ((float) g / 255.0f);
            img[a + 2] = ((float) b / 255.0f);
        }
    }
}

__END_DECLS





<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "layer3x3x64_f.h"

static GLuint _layer3x3x64_f_compute_program_id;

static GLuint _layer3x3x64_f_red_texture_id;
static GLuint _layer3x3x64_f_green_texture_id;
static GLuint _layer3x3x64_f_blue_texture_id;

static GLuint _layer3x3x64_f_output_textures[64];

static GLuint _layer3x3x64_f_kernel_ubo_id;

void layer3x3x64_f_init()
{
    size_t i;

    glGenTextures(64, _layer3x3x64_f_output_textures);
    for (i = 0; i < 64; i++)
    {
        glBindTexture(GL_TEXTURE_2D, _layer3x3x64_f_output_textures[i]);
        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
        glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
        glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);
    }
    glBindTexture(GL_TEXTURE_2D, 0);

    // Image textures allocation
    glGenTextures(1, &_layer3x3x64_f_red_texture_id);
    glGenTextures(1, &_layer3x3x64_f_green_texture_id);
    glGenTextures(1, &_layer3x3x64_f_blue_texture_id);

    glBindTexture(GL_TEXTURE_2D, _layer3x3x64_f_red_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    glBindTexture(GL_TEXTURE_2D, _layer3x3x64_f_green_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    glBindTexture(GL_TEXTURE_2D, _layer3x3x64_f_blue_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_R32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);

    glBindTexture(GL_TEXTURE_2D, 0);

    // Kernels buffer allocation
    glGenBuffers(1, &_layer3x3x64_f_kernel_ubo_id);
    glBindBuffer(GL_UNIFORM_BUFFER, _layer3x3x64_f_kernel_ubo_id);
    glBufferData
        (
            GL_UNIFORM_BUFFER,
            (GLsizeiptr) (9 * 3 * sizeof(GLfloat)),
            NULL,
            GL_DYNAMIC_COPY
        );
    glBindBuffer(GL_UNIFORM_BUFFER, 0);


}

void layer3x3x64_f_update(const float *red, const float *green, const float *blue)
{

}

void layer3x3x64_f_perform_compute()
{

}

void layer3x3x64_f_destroy()
{

}<file_sep>/*
 * Created by <NAME> on 4/27/18 8:38 AM.
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

package limsi.fr.openings;

import android.content.Context;
import android.util.AttributeSet;
import android.util.Log;
import android.view.Surface;
import android.view.SurfaceHolder;
import android.view.SurfaceView;

/**
 * This is an implementation of a custom SurfaceView class that is used by a native code.
 * An EGL context is created with native code and is used to draw in the SurfaceView. In the
 * Java code only initialize the SurfaceView and set callbacks. A native rendering thread is
 * implemented.
 */
public class NativeSurfaceView extends SurfaceView
        implements SurfaceHolder.Callback {

    private final SurfaceHolder _holder;

    /**
     * Basic constructor
     *
     * @param context
     * @param attributes
     */
    public NativeSurfaceView(Context context, AttributeSet attributes) {
        super(context, attributes);

        _holder = getHolder();
        _holder.addCallback(this);
        setWillNotDraw(true); // Important to not draw the SurfaceView here !!
    }

    /**
     * Called when the surface is created, and calls nativeCreateSurface in order to initialize
     * the OpenGLES rendering context.
     *
     * @param surfaceHolder
     */
    @Override
    public void surfaceCreated(SurfaceHolder surfaceHolder) {
        synchronized (_holder) {
            nativeCreateSurface(_holder.getSurface());
        }
        Log.d("NativeSurfaceView", "Surface created");
    }

    /**
     * Called when the surface is resized, and calls nativeResizeSurface to update the OpenGLES
     * context
     *
     * @param surfaceHolder
     * @param format
     * @param width
     * @param height
     */
    @Override
    public void surfaceChanged(SurfaceHolder surfaceHolder, int format, int width, int height) {
        synchronized (_holder) {
            nativeResizeSurface(_holder.getSurface(), format, width, height);
        }
    }

    /**
     * Called when the surface is destroyed and calls nativeDestroySurface that will destroy the
     * created the OpenGLES context
     *
     * @param surfaceHolder
     */
    @Override
    public void surfaceDestroyed(SurfaceHolder surfaceHolder) {
        synchronized (_holder) {
            nativeDestroySurface(_holder.getSurface());
        }
    }

    /*
     * Native methods for the NativeSurfaceView handling
     */
    private static native void nativeCreateSurface(final Surface surface);

    private static native void nativeResizeSurface(final Surface surface, int format,
                                                   int width, int height);

    // private static native void nativeDrawSurface(final Surface surface);

    private static native void nativeDestroySurface(final Surface surface);
}
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "gl_computing.h"

GLuint _output_texture_id;

static float _red_input[COLOR_CAMERA_RESOLUTION];
static float _green_input[COLOR_CAMERA_RESOLUTION];
static float _blue_input[COLOR_CAMERA_RESOLUTION];

// static const float __kernel[] =
//     {
//         1.0f / 16.0f, 1.0f / 8.0f, 1.0f / 16.0f,
//         1.0f / 8.0f, 1.0f / 4.0f, 1.0f / 8.0f,
//         1.0f / 16.0f, 1.0f / 8.0f, 1.0f / 16.0f
//     };

static const float __kernels[] =
    {
        0.0f, -1.0f, 0.0f,
        -1.0f, 4.0f, -1.0f,
        0.0f, -1.0f, 0.0f,
        0.0f, -1.0f, 0.0f,
        -1.0f, 4.0f, -1.0f,
        0.0f, -1.0f, 0.0f,
        0.0f, -1.0f, 0.0f,
        -1.0f, 4.0f, -1.0f,
        0.0f, -1.0f, 0.0f,
    };

// ------------------------------------------------------------------------------------------

static void __print_compute_shader_descr();

// ------------------------------------------------------------------------------------------

void gl_computing_init(void *data)
{
    __print_compute_shader_descr();
    glEnable(GL_TEXTURE_2D);

    glGenTextures(1, &_output_texture_id);
    glBindTexture(GL_TEXTURE_2D, _output_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D(GL_TEXTURE_2D, 1, GL_RGBA32F, COLOR_CAMERA_WIDTH, COLOR_CAMERA_HEIGHT);
    glBindTexture(GL_TEXTURE_2D, 0);

    conv3x3_f_init();
    conv3x3_f_set_kernels(__kernels);
    conv3x3_f_attach_rendering_texture(_output_texture_id);
}

void gl_computing_update(void *data)
{
    struct compute_data *data_ = (struct compute_data *) data;

    size_t i;
    size_t index = 0;

    for (i = 0; i < COLOR_CAMERA_RESOLUTION; i++)
    {
        _red_input[i] = data_->input_image[index++];
        _green_input[i] = data_->input_image[index++];
        _blue_input[i] = data_->input_image[index++];
    }

    conv3x3_f_update(_red_input, _green_input, _blue_input);
}

void gl_computing_compute(void *data)
{
    conv3x3_f_perform_compute();
}

void gl_computing_destroy(void *data)
{
    conv3x3_f_destroy();
}

// ------------------------------------------------------------------------------------------

static void __print_compute_shader_descr()
{
    GLint workgroup_count[3];
    GLint workgroup_size[3];
    GLint workgroup_invocations;
    GLint max_uniform_bock_size;
    GLint uniform_offset_alignment;

    glGetIntegeri_v(GL_MAX_COMPUTE_WORK_GROUP_COUNT, 0, &workgroup_count[0]);
    glGetIntegeri_v(GL_MAX_COMPUTE_WORK_GROUP_COUNT, 1, &workgroup_count[1]);
    glGetIntegeri_v(GL_MAX_COMPUTE_WORK_GROUP_COUNT, 2, &workgroup_count[2]);

    LOGD("Taille maximale des workgroups:\n\tx:%u\n\ty:%u\n\tz:%u\n",
         workgroup_size[0], workgroup_size[1], workgroup_size[2]);

    glGetIntegeri_v(GL_MAX_COMPUTE_WORK_GROUP_SIZE, 0, &workgroup_size[0]);
    glGetIntegeri_v(GL_MAX_COMPUTE_WORK_GROUP_SIZE, 1, &workgroup_size[1]);
    glGetIntegeri_v(GL_MAX_COMPUTE_WORK_GROUP_SIZE, 2, &workgroup_size[2]);
    LOGD("Nombre maximal d'invocations locales:\n\tx:%u\n\ty:%u\n\tz:%u\n",
         workgroup_size[0], workgroup_size[1], workgroup_size[2]);

    glGetIntegerv(GL_MAX_COMPUTE_WORK_GROUP_INVOCATIONS, &workgroup_invocations);
    LOGD("Nombre maximum d'invocations de workgroups:\n\t%u\n", workgroup_invocations);

    glGetIntegerv(GL_MAX_UNIFORM_BLOCK_SIZE, &max_uniform_bock_size);
    LOGD("Taille maximale des UBO : %d", max_uniform_bock_size);

    glGetIntegerv(GL_UNIFORM_BUFFER_OFFSET_ALIGNMENT, &uniform_offset_alignment);
    LOGD("Offset d'alignement : %d", uniform_offset_alignment);
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "model.h"

static struct compute_data _compute_data;
static struct compute_data *_compute_data_ptr;

static struct display_data _display_data;
static struct display_data *_display_data_ptr;

void model_init()
{
    _compute_data_ptr = &_compute_data;
    _display_data_ptr = &_display_data;
}

void model_get_compute_data(struct compute_data **data)
{
    SWAP_POINTERS(&_compute_data_ptr, data);
}

void model_set_compute_data(struct compute_data **data)
{
    SWAP_POINTERS(&_compute_data_ptr, data);
}

void model_get_display_data(struct display_data **data)
{
    SWAP_POINTERS(&_display_data_ptr, data);
}

void model_set_display_data(struct display_data **data)
{
    SWAP_POINTERS(&_display_data_ptr, data);
}

void model_destroy()
{

}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "cl_computing.h"

GLuint _output_texture_id;

// ---------------------------------------------------------------------------------------------

static void __matrix_add(struct compute_info *info, const float *X, const float *Y, float *Z);

// ---------------------------------------------------------------------------------------------

void cl_computing_init(struct compute_info *info, void *data)
{
    // Initialize OpenGL objects
    glEnable(GL_TEXTURE_2D);

    glGenTextures(1, &_output_texture_id);
    glBindTexture(GL_TEXTURE_2D, _output_texture_id);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MIN_FILTER, GL_NEAREST);
    glTexParameteri(GL_TEXTURE_2D, GL_TEXTURE_MAG_FILTER, GL_NEAREST);
    glTexStorage2D
        (
            GL_TEXTURE_2D,
            1,
            GL_RGBA32F,
            COLOR_CAMERA_WIDTH,
            COLOR_CAMERA_HEIGHT
        );
    glBindTexture(GL_TEXTURE_2D, 0);

    grayscale_init(info, _output_texture_id);
}

void cl_computing_update(struct compute_info *info, void *data)
{
    struct compute_data *data_ = (struct compute_data *) data;
    grayscale_set_input_image(data_->input_image);
}

void cl_computing_compute(struct compute_info *info, void *data)
{
    grayscale_compute();
}

void cl_computing_destroy(struct compute_info *info, void *data)
{
    grayscale_destroy();
    glDeleteTextures(1, &_output_texture_id);
}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "gl_31_functions.h"

PFNGLDISPATCHCOMPUTEPROC __glDispatchCompute = NULL;
PFNGLDISPATCHCOMPUTEINDIRECTPROC __glDispatchComputeIndirect = NULL;
PFNGLDRAWARRAYSINDIRECTPROC __glDrawArraysIndirect = NULL;
PFNGLDRAWELEMENTSINDIRECTPROC __glDrawElementsIndirect = NULL;
PFNGLFRAMEBUFFERPARAMETERIPROC __glFramebufferParameteri = NULL;
PFNGLGETFRAMEBUFFERPARAMETERIVPROC __glGetFramebufferParameteri = NULL;
PFNGLGETPROGRAMINTERFACEIVPROC __glGetProgramInterface = NULL;
PFNGLGETPROGRAMRESOURCEINDEXPROC __glGetProgramSourceIndex = NULL;
PFNGLGETPROGRAMRESOURCENAMEPROC __glGetProgramResourceName = NULL;
PFNGLGETPROGRAMRESOURCEIVPROC __glGetProgramResourceiv = NULL;
PFNGLGETPROGRAMRESOURCELOCATIONPROC __glGetProgramResourceLocation = NULL;
PFNGLUSEPROGRAMSTAGESPROC __glUseProgramStages = NULL;
PFNGLACTIVESHADERPROGRAMPROC __glActiveShaderProgram = NULL;
PFNGLCREATESHADERPROGRAMVPROC __glCreateShaderProgram = NULL;
PFNGLBINDPROGRAMPIPELINEPROC __glBindProgramPipeline = NULL;
PFNGLDELETEPROGRAMPIPELINESPROC __glDeleteProgramPipeline = NULL;
PFNGLGENPROGRAMPIPELINESPROC __glGenProgramPipeline = NULL;
PFNGLISPROGRAMPIPELINEPROC __glIsProgramPipeline = NULL;
PFNGLGETPROGRAMPIPELINEIVPROC __glGetProgramPipelineiv = NULL;
PFNGLPROGRAMUNIFORM1IPROC __glProgramUniform1i = NULL;
PFNGLPROGRAMUNIFORM2IPROC __glProgramUniform2i = NULL;
PFNGLPROGRAMUNIFORM3IPROC __glProgramUniform3i = NULL;
PFNGLPROGRAMUNIFORM4IPROC __glProgramUniform4i = NULL;
PFNGLPROGRAMUNIFORM1UIPROC __glProgramUniform1ui = NULL;
PFNGLPROGRAMUNIFORM2UIPROC __glProgramUniform2ui = NULL;
PFNGLPROGRAMUNIFORM3UIPROC __glProgramUniform3ui = NULL;
PFNGLPROGRAMUNIFORM4UIPROC __glProgramUniform4ui = NULL;
PFNGLPROGRAMUNIFORM1FPROC __glProgramUniform1f = NULL;
PFNGLPROGRAMUNIFORM2FPROC __glProgramUniform2f = NULL;
PFNGLPROGRAMUNIFORM3FPROC __glProgramUniform3f = NULL;
PFNGLPROGRAMUNIFORM4FPROC __glProgramUniform4f = NULL;
PFNGLPROGRAMUNIFORM1IVPROC __glProgramUniform1iv = NULL;
PFNGLPROGRAMUNIFORM2IVPROC __glProgramUniform2iv = NULL;
PFNGLPROGRAMUNIFORM3IVPROC __glProgramUniform3iv = NULL;
PFNGLPROGRAMUNIFORM4IVPROC __glProgramUniform4iv = NULL;
PFNGLPROGRAMUNIFORM1UIVPROC __glProgramUniform1uiv = NULL;
PFNGLPROGRAMUNIFORM2UIVPROC __glProgramUniform2uiv = NULL;
PFNGLPROGRAMUNIFORM3UIVPROC __glProgramUniform3uiv = NULL;
PFNGLPROGRAMUNIFORM4UIVPROC __glProgramUniform4uiv = NULL;
PFNGLPROGRAMUNIFORM1FVPROC __glProgramUniform1fv = NULL;
PFNGLPROGRAMUNIFORM2FVPROC __glProgramUniform2fv = NULL;
PFNGLPROGRAMUNIFORM3FVPROC __glProgramUniform3fv = NULL;
PFNGLPROGRAMUNIFORM4FVPROC __glProgramUniform4fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX2FVPROC __glProgramUniformMatrix2fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX3FVPROC __glProgramUniformMatrix3fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX4FVPROC __glProgramUniformMatrix4fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX2X3FVPROC __glProgramUniformMatrix2x3fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX3X2FVPROC __glProgramUniformMatrix3x2fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX2X4FVPROC __glProgramUniformMatrix2x4fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX4X2FVPROC __glProgramUniformMatrix4x2fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX3X4FVPROC __glProgramUniformMatrix3x4fv = NULL;
PFNGLPROGRAMUNIFORMMATRIX4X3FVPROC __glProgramUniformMatrix4x3fv = NULL;
PFNGLVALIDATEPROGRAMPIPELINEPROC __glValidateProgramPipeline = NULL;
PFNGLGETPROGRAMPIPELINEINFOLOGPROC __glGetProgramPipelineInfoLog = NULL;
PFNGLBINDIMAGETEXTUREPROC __glBindImageTexture = NULL;
PFNGLGETBOOLEANI_VPROC __glGetBooleani_v = NULL;
PFNGLMEMORYBARRIERPROC __glMemoryBarrier = NULL;
PFNGLMEMORYBARRIERBYREGIONPROC __glMemoryBarrierByRegion = NULL;
PFNGLTEXSTORAGE2DMULTISAMPLEPROC __glTexStorage2DMultiSample = NULL;
PFNGLGETMULTISAMPLEFVPROC __glGetMultisamplefv = NULL;
PFNGLSAMPLEMASKIPROC __glSampleMaski = NULL;
PFNGLGETTEXLEVELPARAMETERIVPROC __glGetTexLevelParameteriv = NULL;
PFNGLGETTEXLEVELPARAMETERFVPROC __glGetTexLevelParameterfv = NULL;
PFNGLBINDVERTEXBUFFERPROC __glBindVertexBuffer = NULL;
PFNGLVERTEXATTRIBFORMATPROC __glVertexAttribFormat = NULL;
PFNGLVERTEXATTRIBIFORMATPROC __glVertexAttribIFormatProc = NULL;
PFNGLVERTEXATTRIBBINDINGPROC __glVertexAttribBinding = NULL;
PFNGLVERTEXBINDINGDIVISORPROC __glVertexBindingDivisor = NULL;

void gl_functions_load()
{
    __glDispatchCompute
        = (PFNGLDISPATCHCOMPUTEPROC) eglGetProcAddress("glDispatchCompute");
    __glDispatchComputeIndirect
        = (PFNGLDISPATCHCOMPUTEINDIRECTPROC) eglGetProcAddress("glDispatchComputeIndirect");
    __glDrawArraysIndirect
        = (PFNGLDRAWARRAYSINDIRECTPROC) eglGetProcAddress("glDrawArraysIndirect");
    __glDrawElementsIndirect
        = (PFNGLDRAWELEMENTSINDIRECTPROC) eglGetProcAddress("glDrawElementsIndirect");
    __glFramebufferParameteri
        = (PFNGLFRAMEBUFFERPARAMETERIPROC) eglGetProcAddress("glFramebufferParameteri");
    __glGetFramebufferParameteri
        = (PFNGLGETFRAMEBUFFERPARAMETERIVPROC) eglGetProcAddress("glGetFramebufferParameteri");
    __glGetProgramInterface
        = (PFNGLGETPROGRAMINTERFACEIVPROC) eglGetProcAddress("glGetProgramInterface");
    __glGetProgramSourceIndex
        = (PFNGLGETPROGRAMRESOURCEINDEXPROC) eglGetProcAddress("glGetProgramSourceIndex");
    __glGetProgramResourceName
        = (PFNGLGETPROGRAMRESOURCENAMEPROC) eglGetProcAddress("glGetProgramResourceName");
    __glGetProgramResourceiv
        = (PFNGLGETPROGRAMRESOURCEIVPROC) eglGetProcAddress("glGetProgramResourceiv");
    __glGetProgramResourceLocation
        = (PFNGLGETPROGRAMRESOURCELOCATIONPROC) eglGetProcAddress("glGetProgramResourceLocation");
    __glUseProgramStages
        = (PFNGLUSEPROGRAMSTAGESPROC) eglGetProcAddress("glUseProgramStages");
    __glActiveShaderProgram
        = (PFNGLACTIVESHADERPROGRAMPROC) eglGetProcAddress("glActiveShaderProgram");
    __glCreateShaderProgram
        = (PFNGLCREATESHADERPROGRAMVPROC) eglGetProcAddress("glCreateShaderProgram");
    __glBindProgramPipeline
        = (PFNGLBINDPROGRAMPIPELINEPROC) eglGetProcAddress("glBindProgramPipeline");
    __glDeleteProgramPipeline
        = (PFNGLDELETEPROGRAMPIPELINESPROC) eglGetProcAddress("glDeleteProgramPipeline");
    __glGenProgramPipeline
        = (PFNGLGENPROGRAMPIPELINESPROC) eglGetProcAddress("glGenProgramPipeline");
    __glIsProgramPipeline
        = (PFNGLISPROGRAMPIPELINEPROC) eglGetProcAddress("glIsProgramPipeline");
    __glGetProgramPipelineiv
        = (PFNGLGETPROGRAMPIPELINEIVPROC) eglGetProcAddress("glGetProgramPipelineiv");
    __glProgramUniform1i
        = (PFNGLPROGRAMUNIFORM1IPROC) eglGetProcAddress("glProgramUniform1i");
    __glProgramUniform2i
        = (PFNGLPROGRAMUNIFORM2IPROC) eglGetProcAddress("glProgramUniform2i");
    __glProgramUniform3i
        = (PFNGLPROGRAMUNIFORM3IPROC) eglGetProcAddress("glProgramUniform3i");
    __glProgramUniform4i
        = (PFNGLPROGRAMUNIFORM4IPROC) eglGetProcAddress("glProgramUniform4i");
    __glProgramUniform1ui
        = (PFNGLPROGRAMUNIFORM1UIPROC) eglGetProcAddress("glProgramUniform1ui");
    __glProgramUniform2ui
        = (PFNGLPROGRAMUNIFORM2UIPROC) eglGetProcAddress("glProgramUniform2ui");
    __glProgramUniform3ui
        = (PFNGLPROGRAMUNIFORM3UIPROC) eglGetProcAddress("glProgramUniform3ui");
    __glProgramUniform4ui
        = (PFNGLPROGRAMUNIFORM4UIPROC) eglGetProcAddress("glProgramUniform4ui");
    __glProgramUniform1f
        = (PFNGLPROGRAMUNIFORM1FPROC) eglGetProcAddress("glProgramUniform1f");
    __glProgramUniform2f
        = (PFNGLPROGRAMUNIFORM2FPROC) eglGetProcAddress("glProgramUniform2f");
    __glProgramUniform3f
        = (PFNGLPROGRAMUNIFORM3FPROC) eglGetProcAddress("glProgramUniform3f");
    __glProgramUniform4f
        = (PFNGLPROGRAMUNIFORM4FPROC) eglGetProcAddress("glProgramUniform4f");
    __glProgramUniform1iv
        = (PFNGLPROGRAMUNIFORM1IVPROC) eglGetProcAddress("glProgramUniform1iv");
    __glProgramUniform2iv
        = (PFNGLPROGRAMUNIFORM2IVPROC) eglGetProcAddress("glProgramUniform2iv");
    __glProgramUniform3iv
        = (PFNGLPROGRAMUNIFORM3IVPROC) eglGetProcAddress("glProgramUniform3iv");
    __glProgramUniform4iv
        = (PFNGLPROGRAMUNIFORM4IVPROC) eglGetProcAddress("glProgramUniform4iv");
    __glProgramUniform1uiv
        = (PFNGLPROGRAMUNIFORM1UIVPROC) eglGetProcAddress("glProgramUniform1uiv");
    __glProgramUniform2uiv
        = (PFNGLPROGRAMUNIFORM2UIVPROC) eglGetProcAddress("glProgramUniform2uiv");
    __glProgramUniform3uiv
        = (PFNGLPROGRAMUNIFORM3UIVPROC) eglGetProcAddress("glProgramUniform3uiv");
    __glProgramUniform4uiv
        = (PFNGLPROGRAMUNIFORM4UIVPROC) eglGetProcAddress("glProgramUniform4uiv");
    __glProgramUniform1fv
        = (PFNGLPROGRAMUNIFORM1FVPROC) eglGetProcAddress("glProgramUniform1fv");
    __glProgramUniform2fv
        = (PFNGLPROGRAMUNIFORM2FVPROC) eglGetProcAddress("glProgramUniform2fv");
    __glProgramUniform3fv
        = (PFNGLPROGRAMUNIFORM3FVPROC) eglGetProcAddress("glProgramUniform3fv");
    __glProgramUniform4fv
        = (PFNGLPROGRAMUNIFORM4FVPROC) eglGetProcAddress("glProgramUniform4fv");
    __glProgramUniformMatrix2fv
        = (PFNGLPROGRAMUNIFORMMATRIX2FVPROC) eglGetProcAddress("glProgramUniformMatrix2fv");
    __glProgramUniformMatrix3fv
        = (PFNGLPROGRAMUNIFORMMATRIX3FVPROC) eglGetProcAddress("glProgramUniformMatrix3fv");
    __glProgramUniformMatrix4fv
        = (PFNGLPROGRAMUNIFORMMATRIX4FVPROC) eglGetProcAddress("glProgramUniformMatrix4fv");
    __glProgramUniformMatrix2x3fv
        = (PFNGLPROGRAMUNIFORMMATRIX2X3FVPROC) eglGetProcAddress("glProgramUniformMatrix2x3fv");
    __glProgramUniformMatrix3x2fv
        = (PFNGLPROGRAMUNIFORMMATRIX3X2FVPROC) eglGetProcAddress("glProgramUniformMatrix3x2fv");
    __glProgramUniformMatrix2x4fv
        = (PFNGLPROGRAMUNIFORMMATRIX2X4FVPROC) eglGetProcAddress("glProgramUniformMatrix2x4fv");
    __glProgramUniformMatrix4x2fv
        = (PFNGLPROGRAMUNIFORMMATRIX4X2FVPROC) eglGetProcAddress("glProgramUniformMatrix4x2fv");
    __glProgramUniformMatrix3x4fv
        = (PFNGLPROGRAMUNIFORMMATRIX3X4FVPROC) eglGetProcAddress("glProgramUniformMatrix3x4fv");
    __glProgramUniformMatrix4x3fv
        = (PFNGLPROGRAMUNIFORMMATRIX4X3FVPROC) eglGetProcAddress("glProgramUniformMatrix4x3fv");
    __glValidateProgramPipeline
        = (PFNGLVALIDATEPROGRAMPIPELINEPROC) eglGetProcAddress("glValidateProgramPipeline");
    __glGetProgramPipelineInfoLog
        = (PFNGLGETPROGRAMPIPELINEINFOLOGPROC) eglGetProcAddress("glGetProgramPipelineInfoLog");
    __glBindImageTexture
        = (PFNGLBINDIMAGETEXTUREPROC) eglGetProcAddress("glBindImageTexture");
    __glGetBooleani_v
        = (PFNGLGETBOOLEANI_VPROC) eglGetProcAddress("glGetBooleani_v");
    __glMemoryBarrier
        = (PFNGLMEMORYBARRIERPROC) eglGetProcAddress("glMemoryBarrier");
    __glMemoryBarrierByRegion
        = (PFNGLMEMORYBARRIERBYREGIONPROC) eglGetProcAddress("glMemoryBarrierByRegion");
    __glTexStorage2DMultiSample
        = (PFNGLTEXSTORAGE2DMULTISAMPLEPROC) eglGetProcAddress("glTexStorage2DMultiSample");
    __glGetMultisamplefv
        = (PFNGLGETMULTISAMPLEFVPROC) eglGetProcAddress("glGetMultisamplefv");
    __glSampleMaski
        = (PFNGLSAMPLEMASKIPROC) eglGetProcAddress("glSampleMaski");
    __glGetTexLevelParameteriv
        = (PFNGLGETTEXLEVELPARAMETERIVPROC) eglGetProcAddress("glGetTexLevelParameteriv");
    __glGetTexLevelParameterfv
        = (PFNGLGETTEXLEVELPARAMETERFVPROC) eglGetProcAddress("glGetTexLevelParameterfv");
    __glBindVertexBuffer
        = (PFNGLBINDVERTEXBUFFERPROC) eglGetProcAddress("glBindVertexBuffer");
    __glVertexAttribFormat
        = (PFNGLVERTEXATTRIBFORMATPROC) eglGetProcAddress("glVertexAttribFormat");
    __glVertexAttribIFormatProc
        = (PFNGLVERTEXATTRIBIFORMATPROC) eglGetProcAddress("glVertexAttribIFormatPro");
    __glVertexAttribBinding
        = (PFNGLVERTEXATTRIBBINDINGPROC) eglGetProcAddress("glVertexAttribBinding");
    __glVertexBindingDivisor
        = (PFNGLVERTEXBINDINGDIVISORPROC) eglGetProcAddress("glVertexBindingDivisor");

}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_COMPUTING_H
#define OPENINGS_COMPUTING_H

#include <stdint.h>
#include <omp.h>
#include <arm_neon.h>
#include <GLES3/gl3.h>

#include "gl_computing/conv3x3_f.h"
#include "global/global.h"
#include "gl/gl_31_functions.h"
#include "gl/shader.h"
#include "gl/shader_sources.h"
#include "model/model.h"

__BEGIN_DECLS

extern void gl_computing_init(void *data);

extern void gl_computing_update(void *data);

extern void gl_computing_compute(void *data);

extern void gl_computing_destroy(void *data);

__END_DECLS

#endif // OPENINGS_COMPUTING_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef __KERNEL_SOURCES_H
#define __KERNEL_SOURCES_H

static const char *grayscale_kernel_source =
    {
        "/*\n"
        " * Created by <NAME>\n"
        " *\n"
        " * This program is free software: you can redistribute it and/or modify\n"
        " * it under the terms of the GNU General Public License as published by\n"
        " * the Free Software Foundation, either version 3 of the License, or\n"
        " * (at your option) any later version.\n"
        " *\n"
        " * This program is distributed in the hope that it will be useful,\n"
        " * but WITHOUT ANY WARRANTY; without even the implied warranty of\n"
        " * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the\n"
        " * GNU General Public License for more details.\n"
        " *\n"
        " * You should have received a copy of the GNU General Public License\n"
        " */\n"
        "\n"
        "__constant sampler_t _sampler = CLK_NORMALIZED_COORDS_FALSE | CLK_ADDRESS_CLAMP_TO_EDGE | CLK_FILTER_NEAREST;\n"
        "\n"
        "__kernel void img_grayscale\n"
        "    (\n"
        "        __read_only image2d_t input_red_image,\n"
        "        __read_only image2d_t input_green_image,\n"
        "        __read_only image2d_t input_blue_image,\n"
        "        __write_only image2d_t output_image\n"
        "    )\n"
        "{\n"
        "    size_t j = get_global_id(1);\n"
        "    size_t i = get_global_id(0);\n"
        "    float4 red = read_imagef(input_red_image, (int2)(i,j));\n"
        "    float4 green = read_imagef(input_green_image, (int2)(i,j));\n"
        "    float4 blue = read_imagef(input_blue_image, (int2)(i,j));\n"
        "    \n"
        "    float val = (red.x + green.x + blue.x) / 3.0f;\n"
        "    \n"
        "    write_imagef(output_image, (int2)(i,j), (float4)(val, val, val, 1.0f));\n"
        "}\n"
    };

#endif // __KERNEL_SOURCES_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "shader.h"

GLuint shader_createProgram(const char *vtxSrc, const char *fragSrc)
{
    GLuint vtxShader = 0;
    GLuint fragShader = 0;
    GLuint program = 0;
    GLint linked = GL_FALSE;

    vtxShader = shader_createShader(GL_VERTEX_SHADER, vtxSrc);
    if (!vtxShader)
        goto exit;

    fragShader = shader_createShader(GL_FRAGMENT_SHADER, fragSrc);
    if (!fragShader)
        goto exit;

    program = glCreateProgram();
    if (!program)
    {
        //checkGlError("glCreateProgram");
        goto exit;
    }
    glAttachShader(program, vtxShader);
    glAttachShader(program, fragShader);

    glLinkProgram(program);
    glGetProgramiv(program, GL_LINK_STATUS, &linked);
    if (!linked)
    {
        LOGE("Could not link program");
        GLint infoLogLen = 0;
        glGetProgramiv(program, GL_INFO_LOG_LENGTH, &infoLogLen);
        if (infoLogLen)
        {
            GLchar *infoLog = (GLchar *) malloc(infoLogLen);
            if (infoLog)
            {
                glGetProgramInfoLog(program, infoLogLen, NULL, infoLog);
                LOGE("Could not link program:\n%s\n", infoLog);
                free(infoLog);
            }
        }
        glDeleteProgram(program);
        program = 0;
    }

    exit:
    glDeleteShader(vtxShader);
    glDeleteShader(fragShader);
    return program;
}

GLuint shader_create_compute_program(const char *kernel)
{
    GLuint program_id = 0;
    GLuint compute_shader = 0;
    GLint linked = GL_FALSE;

    compute_shader = shader_createShader(GL_COMPUTE_SHADER, kernel);
    if (!compute_shader) goto exit;

    program_id = glCreateProgram();
    if (!program_id) goto exit;

    glAttachShader(program_id, compute_shader);

    glLinkProgram(program_id);
    glGetProgramiv(program_id, GL_LINK_STATUS, &linked);
    if (!linked)
    {
        LOGE("Could not link compute program");
        GLint infoLogLen = 0;
        glGetProgramiv(program_id, GL_INFO_LOG_LENGTH, &infoLogLen);
        if (infoLogLen)
        {
            GLchar *infoLog = (GLchar *) malloc((size_t) infoLogLen);
            if (infoLog)
            {
                glGetProgramInfoLog(program_id, infoLogLen, NULL, infoLog);
                LOGE("Could not link compute program:\n%s\n", infoLog);
                free(infoLog);
            }
        }
        glDeleteProgram(program_id);
        program_id = 0;
    }
    exit:
    glDeleteShader(compute_shader);
    return program_id;
}

GLuint shader_createShader(GLenum shaderType, const char *src)
{
    GLuint shader = glCreateShader(shaderType);
    if (!shader)
    {
        LOGE("Problem shader");
        return 0;
    }
    glShaderSource(shader, 1, &src, NULL);

    GLint compiled = GL_FALSE;
    glCompileShader(shader);
    glGetShaderiv(shader, GL_COMPILE_STATUS, &compiled);
    if (!compiled)
    {
        GLint infoLogLen = 0;
        glGetShaderiv(shader, GL_INFO_LOG_LENGTH, &infoLogLen);
        if (infoLogLen > 0)
        {
            GLchar *infoLog = (GLchar *) malloc(infoLogLen);
            if (infoLog)
            {
                glGetShaderInfoLog(shader, infoLogLen, NULL, infoLog);

                LOGE
                (
                    "Could not compile %s shader:\n%s\n",
                    shaderType == GL_VERTEX_SHADER
                    ? "vertex"
                    : (shaderType == GL_FRAGMENT_SHADER
                       ? "fragment"
                       : "compute"),
                    infoLog
                );

                free(infoLog);
            }
        }
        glDeleteShader(shader);
        return 0;
    }
    return shader;
}

<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef __SHADER_SOURCES_H
#define __SHADER_SOURCES_H

#include <GLES3/gl31.h>

static const GLchar *CONV3X3_source =
    {
        "/*\n"
        " * Created by <NAME>\n"
        " *\n"
        " * This program is free software: you can redistribute it and/or modify\n"
        " * it under the terms of the GNU General Public License as published by\n"
        " * the Free Software Foundation, either version 3 of the License, or\n"
        " * (at your option) any later version.\n"
        " *\n"
        " * This program is distributed in the hope that it will be useful,\n"
        " * but WITHOUT ANY WARRANTY; without even the implied warranty of\n"
        " * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the\n"
        " * GNU General Public License for more details.\n"
        " *\n"
        " * You should have received a copy of the GNU General Public License\n"
        " */\n"
        "\n"
        "#version 310 es\n"
        "\n"
        "precision highp float;\n"
        "\n"
        "layout(local_size_x = 1, local_size_y = 540, local_size_z = 1) in;\n"
        "\n"
        "// --------------------------------------------------------------------------------------------\n"
        "// ----------------------------------- Kernels ------------------------------------------------\n"
        "// --------------------------------------------------------------------------------------------\n"
        "\n"
        "layout(std140, binding = 0) uniform RED_KERNEL\n"
        "{\n"
        "    float red_kernel[9];\n"
        "};\n"
        "\n"
        "layout(std140, binding = 1) uniform GREEN_KERNEL\n"
        "{\n"
        "    float green_kernel[9];\n"
        "};\n"
        "\n"
        "layout(std140, binding = 2) uniform BLUE_KERNEL\n"
        "{\n"
        "    float blue_kernel[9];\n"
        "};\n"
        "\n"
        "// --------------------------------------------------------------------------------------------\n"
        "// -------------------------------- Input images ----------------------------------------------\n"
        "// --------------------------------------------------------------------------------------------\n"
        "\n"
        "layout(r32f, binding = 3) readonly uniform highp image2D _red_input;\n"
        "layout(r32f, binding = 4) readonly uniform highp image2D _green_input;\n"
        "layout(r32f, binding = 5) readonly uniform highp image2D _blue_input;\n"
        "\n"
        "// --------------------------------------------------------------------------------------------\n"
        "// -------------------------------- Output image ----------------------------------------------\n"
        "// --------------------------------------------------------------------------------------------\n"
        "\n"
        "layout(rgba32f, binding = 6) writeonly uniform highp image2D _output;\n"
        "\n"
        "// -------------------------------------------------------------------------------------------\n"
        "\n"
        "void main()\n"
        "{\n"
        "    uint i = gl_WorkGroupID.x * gl_WorkGroupSize.x + gl_LocalInvocationID.x;\n"
        "    uint j = gl_WorkGroupID.y * gl_WorkGroupSize.y + gl_LocalInvocationID.y;\n"
        "\n"
        "    float val_r;\n"
        "    float val_g;\n"
        "    float val_b;\n"
        "\n"
        "    val_r =   imageLoad(_red_input, ivec2(i - 1u, j - 1u)).r * red_kernel[0]\n"
        "            + imageLoad(_red_input, ivec2(i - 1u, j     )).r * red_kernel[1]\n"
        "            + imageLoad(_red_input, ivec2(i - 1u, j + 1u)).r * red_kernel[2]\n"
        "            + imageLoad(_red_input, ivec2(i     , j - 1u)).r * red_kernel[3]\n"
        "            + imageLoad(_red_input, ivec2(i     , j     )).r * red_kernel[4]\n"
        "            + imageLoad(_red_input, ivec2(i     , j + 1u)).r * red_kernel[5]\n"
        "            + imageLoad(_red_input, ivec2(i + 1u, j - 1u)).r * red_kernel[6]\n"
        "            + imageLoad(_red_input, ivec2(i + 1u, j     )).r * red_kernel[7]\n"
        "            + imageLoad(_red_input, ivec2(i + 1u, j + 1u)).r * red_kernel[8];\n"
        "\n"
        "    val_g =   imageLoad(_green_input, ivec2(i - 1u, j - 1u)).r * green_kernel[0]\n"
        "            + imageLoad(_green_input, ivec2(i - 1u, j     )).r * green_kernel[1]\n"
        "            + imageLoad(_green_input, ivec2(i - 1u, j + 1u)).r * green_kernel[2]\n"
        "            + imageLoad(_green_input, ivec2(i     , j - 1u)).r * green_kernel[3]\n"
        "            + imageLoad(_green_input, ivec2(i     , j     )).r * green_kernel[4]\n"
        "            + imageLoad(_green_input, ivec2(i     , j + 1u)).r * green_kernel[5]\n"
        "            + imageLoad(_green_input, ivec2(i + 1u, j - 1u)).r * green_kernel[6]\n"
        "            + imageLoad(_green_input, ivec2(i + 1u, j     )).r * green_kernel[7]\n"
        "            + imageLoad(_green_input, ivec2(i + 1u, j + 1u)).r * green_kernel[8];\n"
        "\n"
        "    val_b =   imageLoad(_blue_input, ivec2(i - 1u, j - 1u)).r * blue_kernel[0]\n"
        "            + imageLoad(_blue_input, ivec2(i - 1u, j     )).r * blue_kernel[1]\n"
        "            + imageLoad(_blue_input, ivec2(i - 1u, j + 1u)).r * blue_kernel[2]\n"
        "            + imageLoad(_blue_input, ivec2(i     , j - 1u)).r * blue_kernel[3]\n"
        "            + imageLoad(_blue_input, ivec2(i     , j     )).r * blue_kernel[4]\n"
        "            + imageLoad(_blue_input, ivec2(i     , j + 1u)).r * blue_kernel[5]\n"
        "            + imageLoad(_blue_input, ivec2(i + 1u, j - 1u)).r * blue_kernel[6]\n"
        "            + imageLoad(_blue_input, ivec2(i + 1u, j     )).r * blue_kernel[7]\n"
        "            + imageLoad(_blue_input, ivec2(i + 1u, j + 1u)).r * blue_kernel[8];\n"
        "\n"
        "    imageStore(_output, ivec2(i, j), vec4(val_r, val_g, val_b, 1.0f));\n"
        "}\n"
    };

static const GLchar *GRAYSCALE_source =
    {
        "/*\n"
        " * Created by <NAME>\n"
        " *\n"
        " * This program is free software: you can redistribute it and/or modify\n"
        " * it under the terms of the GNU General Public License as published by\n"
        " * the Free Software Foundation, either version 3 of the License, or\n"
        " * (at your option) any later version.\n"
        " *\n"
        " * This program is distributed in the hope that it will be useful,\n"
        " * but WITHOUT ANY WARRANTY; without even the implied warranty of\n"
        " * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the\n"
        " * GNU General Public License for more details.\n"
        " *\n"
        " * You should have received a copy of the GNU General Public License\n"
        " */\n"
        "\n"
        "#version 310 es\n"
        "\n"
        "precision lowp float;\n"
        "\n"
        "layout(local_size_x = 1, local_size_y = 540, local_size_z = 1) in;\n"
        "\n"
        "// --------------------------------------------------------------------------------------------\n"
        "// -------------------------------- Input images ----------------------------------------------\n"
        "// --------------------------------------------------------------------------------------------\n"
        "\n"
        "layout(r32f, binding = 0) readonly uniform highp image2D _red_input;\n"
        "layout(r32f, binding = 1) readonly uniform highp image2D _green_input;\n"
        "layout(r32f, binding = 2) readonly uniform highp image2D _blue_input;\n"
        "\n"
        "// --------------------------------------------------------------------------------------------\n"
        "// -------------------------------- Output image ----------------------------------------------\n"
        "// --------------------------------------------------------------------------------------------\n"
        "\n"
        "layout(rgba32f, binding = 3) writeonly highp uniform image2D _output;\n"
        "\n"
        "// -------------------------------------------------------------------------------------------\n"
        "\n"
        "void main()\n"
        "{\n"
        "    uint i = gl_WorkGroupID.x;\n"
        "    uint j = gl_WorkGroupID.y * gl_WorkGroupSize.y + gl_LocalInvocationID.y;\n"
        "\n"
        "    float r, g, b;\n"
        "    float val;\n"
        "\n"
        "    r = imageLoad(_red_input, ivec2(i, j)).r;\n"
        "    g = imageLoad(_green_input, ivec2(i, j)).r;\n"
        "    b = imageLoad(_blue_input, ivec2(i, j)).r;\n"
        "\n"
        "    val = (r + g + b) / 3.0f;\n"
        "\n"
        "    imageStore(_output, ivec2(i, j), vec4(val, val, val, 1.0f));\n"
        "}\n"
    };

static const GLchar *MATRIX_SAXPY_source = 
	{
		"/*\n"
		" * Created by <NAME>\n"
		" *\n"
		" * This program is free software: you can redistribute it and/or modify\n"
		" * it under the terms of the GNU General Public License as published by\n"
		" * the Free Software Foundation, either version 3 of the License, or\n"
		" * (at your option) any later version.\n"
		" *\n"
		" * This program is distributed in the hope that it will be useful,\n"
		" * but WITHOUT ANY WARRANTY; without even the implied warranty of\n"
		" * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the\n"
		" * GNU General Public License for more details.\n"
		" *\n"
		" * You should have received a copy of the GNU General Public License\n"
		" */\n"
		"\n"
		"#version 310 es\n"
		"\n"
		"layout(local_size_x = 1024, local_size_y = 1, local_size_z = 1) in;\n"
		"\n"
		"layout(std430, binding = 0) readonly buffer A_buffer\n"
		"{\n"
		"    float data[];\n"
		"} A;\n"
		"\n"
		"layout(std430, binding = 1) readonly buffer X_buffer\n"
		"{\n"
		"    float data[];\n"
		"} X;\n"
		"\n"
		"layout(std430, binding = 2) readonly buffer Y_buffer\n"
		"{\n"
		"    float data[];\n"
		"} Y;\n"
		"\n"
		"layout(std430, binding = 3) buffer Z_buffer\n"
		"{\n"
		"    float data[];\n"
		"} Z;\n"
		"\n"
		"void main()\n"
		"{\n"
		"    uint i = gl_WorkGroupID.x;\n"
		"    uint j = gl_LocalInvocationID.x;\n"
		"    uint num_cols = gl_NumWorkGroups.x;\n"
		"\n"
		"    uint Z_index = i * num_cols + j;\n"
		"    uint A_index = i * num_cols;\n"
		"\n"
		"    Z.data[Z_index] = 0.0f;\n"
		"\n"
		"    for(uint id = A_index; id < (A_index + num_cols); id++)\n"
		"    {\n"
		"        Z.data[Z_index] += A.data[id] * X.data[id];\n"
		"    }\n"
		"    Z.data[Z_index] += Y.data[Z_index];\n"
		"}\n"
	};

static const GLchar *SAXPY_source = 
	{
		"/*\n"
		" * Created by <NAME>\n"
		" *\n"
		" * This program is free software: you can redistribute it and/or modify\n"
		" * it under the terms of the GNU General Public License as published by\n"
		" * the Free Software Foundation, either version 3 of the License, or\n"
		" * (at your option) any later version.\n"
		" *\n"
		" * This program is distributed in the hope that it will be useful,\n"
		" * but WITHOUT ANY WARRANTY; without even the implied warranty of\n"
		" * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the\n"
		" * GNU General Public License for more details.\n"
		" *\n"
		" * You should have received a copy of the GNU General Public License\n"
		" */\n"
		"\n"
		"#version 310 es\n"
		"\n"
		"// Performs a single SAXPY operation\n"
		"\n"
		"layout(local_size_x = 1280) in;\n"
		"\n"
		"layout(std430, binding = 0) readonly buffer X_buffer\n"
		"{\n"
		"    float elements[];\n"
		"} X;\n"
		"\n"
		"layout(std430, binding = 1) readonly buffer Y_buffer\n"
		"{\n"
		"    float elements[];\n"
		"} Y;\n"
		"\n"
		"layout(std430, binding = 2) writeonly buffer Z_buffer\n"
		"{\n"
		"    float elements[];\n"
		"} Z;\n"
		"\n"
		"uniform float SAXPY_A;\n"
		"\n"
		"void main()\n"
		"{\n"
		"    Z.elements[gl_GlobalInvocationID.x]\n"
		"        = SAXPY_A * X.elements[gl_GlobalInvocationID.x] + Y.elements[gl_GlobalInvocationID.x];\n"
		"}\n"
	};

#endif // __SHADER_SOURCES_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_LAYER3X3X32_F_H
#define OPENINGS_LAYER3X3X32_F_H

#include <stdint.h>
#include <GLES3/gl3.h>

#include "global/global.h"
#include "gl/gl_31_functions.h"
#include "gl/shader.h"
#include "gl/shader_sources.h"

__BEGIN_DECLS

extern void layer3x3x64_f_init();

extern void layer3x3x64_f_update(const float *red, const float *green, const float *blue);

extern void layer3x3x64_f_perform_compute();

extern void layer3x3x64_f_destroy();

__END_DECLS

#endif // OPENINGS_LAYER3X3X32_F_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "rendering.h"

static const GLchar _vertex_shader_source[] =
    "#version 310 es\n"
        "precision mediump float;\n"
        "layout(location = 0) in vec3 _position;\n"
        "layout(location = 1) in vec2 _texcoords;\n"
        "out vec2 uv;\n"
        "void main(){\n"
        "gl_Position = vec4(_position.y, -_position.x, _position.z, 1.0);\n"
        "uv = vec2(_texcoords);\n"
        "}\n";

static const GLchar _fragment_shader_source[] =
    "#version 310 es\n"
        "precision mediump float;\n"
        "precision mediump sampler2D;\n"
        "in vec2 uv;\n"
        "uniform sampler2D tex;\n"
        "layout(location = 0) out vec4 out_color;\n"
        "void main(){\n"
        "out_color = texture(tex, uv);\n"
        "}\n";

static const GLfloat _positions[] =
    {
        -1.0f, -1.0f, 0.0f,
        1.0f, -1.0f, 0.0f,
        -1.0f, 1.0f, 0.0f,
        1.0f, 1.0f, 0.0f,
        1.0f, -1.0f, 0.0f,
        -1.0f, 1.0f, 0.0f
    };

static const GLfloat _texcoords[] =
    {
        0.0f, 1.0f,
        1.0f, 1.0f,
        0.0f, 0.0f,
        1.0f, 0.0f,
        1.0f, 1.0f,
        0.0f, 0.0f
    };

static GLuint _program;
static GLuint _positions_vbo;
static GLuint _texcoords_vbo;

extern GLuint _output_texture_id;

// ----------------------------------------------------------------------------------------


void rendering_init(struct display_data *data)
{
    glEnable(GL_TEXTURE_2D);
    glClearColor(1.0f, 0.0f, 0.0f, 1.0f);
    _program = shader_createProgram(_vertex_shader_source, _fragment_shader_source);

    glGenBuffers(1, &_positions_vbo);
    glGenBuffers(1, &_texcoords_vbo);

    glBindBuffer(GL_ARRAY_BUFFER, _positions_vbo);
    glBufferData(GL_ARRAY_BUFFER, 18 * sizeof(float), _positions, GL_STATIC_DRAW);

    glBindBuffer(GL_ARRAY_BUFFER, _texcoords_vbo);
    glBufferData(GL_ARRAY_BUFFER, 12 * sizeof(float), _texcoords, GL_STATIC_DRAW);

    glBindBuffer(GL_ARRAY_BUFFER, 0);
}

void rendering_resize(struct display_data *data, GLsizei w, GLsizei h)
{
    glViewport(0, 0, w, h);
}

void rendering_update(struct display_data *data)
{

}

void rendering_clear(struct display_data *data)
{

}

void rendering_draw(struct display_data *data)
{
    glClear(GL_COLOR_BUFFER_BIT);

    glUseProgram(_program);

    glActiveTexture(GL_TEXTURE0);
    glBindTexture(GL_TEXTURE_2D, _output_texture_id);

    glBindBuffer(GL_ARRAY_BUFFER, _positions_vbo);
    glEnableVertexAttribArray(0);
    glVertexAttribPointer(0, 3, GL_FLOAT, GL_FALSE, 0, 0);

    glBindBuffer(GL_ARRAY_BUFFER, _texcoords_vbo);
    glEnableVertexAttribArray(1);
    glVertexAttribPointer(1, 2, GL_FLOAT, GL_FALSE, 0, 0);

    glDrawArrays(GL_TRIANGLES, 0, 6);

    glUseProgram(0);

    glFlush();
}

void rendering_destroy(struct display_data *data)
{

}<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#ifndef OPENINGS_MODEL_H
#define OPENINGS_MODEL_H

#include <GLES3/gl3.h>

#include"global/global.h"

#if (USE_OPENCL == 1)
struct compute_info
{
    cl_platform_id platform_id;
    cl_device_id device_id;
    cl_context context;
    cl_command_queue command_queue;
};
#endif

struct compute_data
{
    float input_image[COLOR_CAMERA_IMAGE_SIZE];
};

struct display_data
{
    // uint8_t output_image[COLOR_CAMERA_IMAGE_SIZE];
    GLuint _output_id;
};

__BEGIN_DECLS

extern void model_init();

extern void model_get_compute_data(struct compute_data **data);

extern void model_set_compute_data(struct compute_data **data);

extern void model_get_display_data(struct display_data **data);

extern void model_set_display_data(struct display_data **data);

extern void model_destroy();

__END_DECLS

#endif // OPENINGS_MODEL_H
<file_sep>/*
 * Created by <NAME>
 *
 * This program is free software: you can redistribute it and/or modify
 * it under the terms of the GNU General Public License as published by
 * the Free Software Foundation, either version 3 of the License, or
 * (at your option) any later version.
 *
 * This program is distributed in the hope that it will be useful,
 * but WITHOUT ANY WARRANTY; without even the implied warranty of
 * MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
 * GNU General Public License for more details.
 *
 * You should have received a copy of the GNU General Public License
 */

#include "cl_context.h"

static struct compute_info _info;

void cl_context_init_computing(struct compute_data *data)
{
    cl_int err;
    cl_uint num_platforms;
    cl_uint num_devices;

    _info.platform_id = NULL;

    err = clGetPlatformIDs(1, &(_info.platform_id), &num_platforms);
    checkCLError("getPlatformIDs", err);
    if (err != CL_SUCCESS)
    {
        return;
    }

    err = clGetDeviceIDs(_info.platform_id, CL_DEVICE_TYPE_GPU, 1, &(_info.device_id),
                         &num_devices);
    checkCLError("getDevicesIDs", err);
    if (err != CL_SUCCESS)
    {
        return;
    }
    LOGD("There are : %d devices available", num_devices);

    cl_context_properties props[] =
        {
            CL_GL_CONTEXT_KHR, (cl_context_properties) eglGetCurrentContext(),
            CL_EGL_DISPLAY_KHR, (cl_context_properties) eglGetCurrentDisplay(),
            CL_CONTEXT_PLATFORM, (cl_context_properties) _info.platform_id,
            CL_NONE
        };

    _info.context = clCreateContextFromType(props, CL_DEVICE_TYPE_GPU, NULL, NULL, &err);
    checkCLError("Creating an OpenCL context", err);
    if (err != CL_SUCCESS)
    {
        return;
    }

    _info.command_queue =
        clCreateCommandQueueWithProperties(_info.context, _info.device_id, NULL, &err);
    checkCLError("Creating command queue", err);
    if (err != CL_SUCCESS)
    {
        return;
    }

    LOGD("Context created");

    cl_computing_init(&_info, data);
}

void cl_context_update_computing(struct compute_data *data)
{
    cl_computing_update(&_info, data);
}

void cl_context_perform_computing(struct compute_data *data)
{
    cl_computing_compute(&_info, data);
}

void cl_context_destroy_computing(struct compute_data *data)
{
    cl_computing_destroy(&_info, data);

    clReleaseCommandQueue(_info.command_queue);
    clReleaseContext(_info.context);
    clReleaseDevice(_info.device_id);
}<file_sep># Created by <NAME>
#
# This program is free software: you can redistribute it and/or modify
# it under the terms of the GNU General Public License as published by
# the Free Software Foundation, either version 3 of the License, or
# (at your option) any later version.
#
# This program is distributed in the hope that it will be useful,
# but WITHOUT ANY WARRANTY; without even the implied warranty of
# MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
# GNU General Public License for more details.
#
# You should have received a copy of the GNU General Public License

cmake_minimum_required(VERSION 3.6)

set(CMAKE_C_STANDARD 11)
set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} -DUSE_OPENCL=1")
set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} -std=c11 -O3 -fopenmp")
set(CMAKE_C_FLAGS "${CMAKE_C_FLAGS} ${OpenMP_C_FLAGS}")
set(CMAKE_EXE_LINKER_FLAGS "${CMAKE_EXE_LINKER_FLAGS} ${OpenMP_EXE_LINKER_FLAGS}")

## --------------------------------------------------------------------------------------------
## --------------------- List all include files -----------------------------------------------
## --------------------------------------------------------------------------------------------

include_directories(
    ${CMAKE_CURRENT_SOURCE_DIR}/
    ${CMAKE_CURRENT_SOURCE_DIR}/application
    ${CMAKE_CURRENT_SOURCE_DIR}/CL
    ${CMAKE_CURRENT_SOURCE_DIR}/cl_computing
    ${CMAKE_CURRENT_SOURCE_DIR}/cl_context
    ${CMAKE_CURRENT_SOURCE_DIR}/gl
    ${CMAKE_CURRENT_SOURCE_DIR}/gl_computing
    ${CMAKE_CURRENT_SOURCE_DIR}/global
    ${CMAKE_CURRENT_SOURCE_DIR}/model
    ${CMAKE_CURRENT_SOURCE_DIR}/rendering
)

## --------------------------------------------------------------------------------------------
## ---------------------- List all src files --------------------------------------------------
## --------------------------------------------------------------------------------------------

set(
    APPLICATION_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/application/application.c
)

set(
    CL_CONTEXT_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/cl_context/cl_context.c
    ${CMAKE_CURRENT_SOURCE_DIR}/cl_context/kernel.c
)

set(
    CL_COMPUTING_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/cl_computing/cl_computing.c
        ${CMAKE_CURRENT_SOURCE_DIR}/cl_computing/cl_grayscale.c
)

set(
    GL_COMPUTING_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/gl_computing/gl_computing.c
    ${CMAKE_CURRENT_SOURCE_DIR}/gl_computing/conv3x3_f.c
    ${CMAKE_CURRENT_SOURCE_DIR}/gl_computing/grayscale_f.c
    ${CMAKE_CURRENT_SOURCE_DIR}/gl_computing/layer3x3x64_f.c
)

set(
    GL_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/gl/gl_31_functions.c
    ${CMAKE_CURRENT_SOURCE_DIR}/gl/gl_context.c
    ${CMAKE_CURRENT_SOURCE_DIR}/gl/gl_thread.c
    ${CMAKE_CURRENT_SOURCE_DIR}/gl/shader.c
)

set(
    GLOBAL_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/global/global.c
)

set(
    MODEL_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/model/model.c
)

set(
    RENDERING_SRCS
    ${CMAKE_CURRENT_SOURCE_DIR}/rendering/rendering.c
)

set(
    SRC_FILES
    ${CMAKE_CURRENT_SOURCE_DIR}/native_interface.c
    ${APPLICATION_SRCS}
    ${CL_CONTEXT_SRCS}
    ${CL_COMPUTING_SRCS}
    ${GL_COMPUTING_SRCS}
    ${GL_SRCS}
    ${GLOBAL_SRCS}
    ${MODEL_SRCS}
    ${RENDERING_SRCS}
)

## --------------------------------------------------------------------------------------------
## ---------------------------- List libraries ------------------------------------------------
## --------------------------------------------------------------------------------------------

add_library(
    OpenCL
    SHARED
    IMPORTED
)

set_target_properties(
    OpenCL
    PROPERTIES IMPORTED_LOCATION
    ${CMAKE_CURRENT_SOURCE_DIR}/../jniLibs/libOpenCL.so
)

## --------------------------------------------------------------------------------------------
## --------------------------------- Build ----------------------------------------------------
## --------------------------------------------------------------------------------------------

add_library(native_interface SHARED ${SRC_FILES})
target_link_libraries(
    native_interface
    OpenCL
    android
    EGL
    GLESv3
    log
    m
)