tmaq
/
tornavis
1
0
Fork 0
Code Issues 2 Pull Requests 29 Packages Projects 24 Releases Wiki Activity

Cleanup: Remove now-unused "tiled" compositor implementation 

New ("fullframe") CPU compositor backend is being used now, and all the code
related to "tiled" CPU compositor is just never used anymore. The new backend
is faster, uses less memory, better matches GPU compositor, etc.

TL;DR: 20 thousand lines of code gone.

This commit:
- Removes various bits and pieces related to "tiled" compositor (execution
  groups, one-pixel-at-a-time node processing, read/write buffer operations
  related to node execution groups).
- "GPU" (OpenCL) execution device, that was only used by several nodes of
  the tiled compositor.
  - With that, remove CLEW external library too, since nothing within Blender
    uses OpenCL directly anymore.

Pull Request: https://projects.blender.org/blender/blender/pulls/118819
This commit is contained in:
Aras Pranckevicius 2024-02-28 16:59:16 +01:00 committed by Aras Pranckevicius
parent d296c66ab1
commit f5f7024040
296 changed files with 256 additions and 20633 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

3
extern/CMakeLists.txt vendored
View File

@ -55,8 +55,7 @@ if(WITH_LZMA)
add_subdirectory(lzma)
endif()
if(WITH_CYCLES OR WITH_COMPOSITOR_CPU OR WITH_OPENSUBDIV)
add_subdirectory(clew)
if(WITH_CYCLES OR WITH_OPENSUBDIV)
if((WITH_CYCLES_DEVICE_CUDA OR WITH_CYCLES_DEVICE_OPTIX) AND WITH_CUDA_DYNLOAD)
add_subdirectory(cuew)
endif()

24
extern/clew/CMakeLists.txt vendored
View File

@ -1,24 +0,0 @@
# SPDX-FileCopyrightText: 2006 Blender Foundation
#
# SPDX-License-Identifier: GPL-2.0-or-later
set(INC
.
include
)
set(INC_SYS
)
set(SRC
include/clew.h
src/clew.c
)
set(LIB
)
add_definitions(-DCL_USE_DEPRECATED_OPENCL_1_1_APIS)
blender_add_lib(extern_clew "${SRC}" "${INC}" "${INC_SYS}" "${LIB}")

5
extern/clew/README.blender vendored
View File

@ -1,5 +0,0 @@
Project: OpenCL Wrangler
URL: https://github.com/OpenCLWrangler/clew
License: Apache 2.0
Upstream version: 27a6867
Local modifications: None

2822
extern/clew/include/clew.h vendored
View File

File diff suppressed because it is too large Load Diff

407
extern/clew/src/clew.c vendored
View File

@ -1,407 +0,0 @@
//////////////////////////////////////////////////////////////////////////
// Copyright (c) 2009 Organic Vectory B.V.
// Written by George van Venrooij
//
// Distributed under the Boost Software License, Version 1.0.
// (See accompanying file license.txt)
//////////////////////////////////////////////////////////////////////////
#include "clew.h"
#ifdef _WIN32
#define WIN32_LEAN_AND_MEAN
#define VC_EXTRALEAN
#include <windows.h>
typedef HMODULE CLEW_DYNLIB_HANDLE;
#define CLEW_DYNLIB_OPEN LoadLibraryA
#define CLEW_DYNLIB_CLOSE FreeLibrary
#define CLEW_DYNLIB_IMPORT GetProcAddress
#else
#include <dlfcn.h>
typedef void* CLEW_DYNLIB_HANDLE;
#define CLEW_DYNLIB_OPEN(path) dlopen(path, RTLD_NOW | RTLD_GLOBAL)
#define CLEW_DYNLIB_CLOSE dlclose
#define CLEW_DYNLIB_IMPORT dlsym
#endif
#include <stdlib.h>
//! \brief module handle
static CLEW_DYNLIB_HANDLE module = NULL;
// Variables holding function entry points
PFNCLGETPLATFORMIDS __clewGetPlatformIDs = NULL;
PFNCLGETPLATFORMINFO __clewGetPlatformInfo = NULL;
PFNCLGETDEVICEIDS __clewGetDeviceIDs = NULL;
PFNCLGETDEVICEINFO __clewGetDeviceInfo = NULL;
PFNCLCREATESUBDEVICES __clewCreateSubDevices = NULL;
PFNCLRETAINDEVICE __clewRetainDevice = NULL;
PFNCLRELEASEDEVICE __clewReleaseDevice = NULL;
PFNCLCREATECONTEXT __clewCreateContext = NULL;
PFNCLCREATECONTEXTFROMTYPE __clewCreateContextFromType = NULL;
PFNCLRETAINCONTEXT __clewRetainContext = NULL;
PFNCLRELEASECONTEXT __clewReleaseContext = NULL;
PFNCLGETCONTEXTINFO __clewGetContextInfo = NULL;
PFNCLCREATECOMMANDQUEUE __clewCreateCommandQueue = NULL;
PFNCLRETAINCOMMANDQUEUE __clewRetainCommandQueue = NULL;
PFNCLRELEASECOMMANDQUEUE __clewReleaseCommandQueue = NULL;
PFNCLGETCOMMANDQUEUEINFO __clewGetCommandQueueInfo = NULL;
#ifdef CL_USE_DEPRECATED_OPENCL_1_0_APIS
PFNCLSETCOMMANDQUEUEPROPERTY __clewSetCommandQueueProperty = NULL;
#endif
PFNCLCREATEBUFFER __clewCreateBuffer = NULL;
PFNCLCREATESUBBUFFER __clewCreateSubBuffer = NULL;
PFNCLCREATEIMAGE __clewCreateImage = NULL;
PFNCLRETAINMEMOBJECT __clewRetainMemObject = NULL;
PFNCLRELEASEMEMOBJECT __clewReleaseMemObject = NULL;
PFNCLGETSUPPORTEDIMAGEFORMATS __clewGetSupportedImageFormats = NULL;
PFNCLGETMEMOBJECTINFO __clewGetMemObjectInfo = NULL;
PFNCLGETIMAGEINFO __clewGetImageInfo = NULL;
PFNCLSETMEMOBJECTDESTRUCTORCALLBACK __clewSetMemObjectDestructorCallback = NULL;
PFNCLCREATESAMPLER __clewCreateSampler = NULL;
PFNCLRETAINSAMPLER __clewRetainSampler = NULL;
PFNCLRELEASESAMPLER __clewReleaseSampler = NULL;
PFNCLGETSAMPLERINFO __clewGetSamplerInfo = NULL;
PFNCLCREATEPROGRAMWITHSOURCE __clewCreateProgramWithSource = NULL;
PFNCLCREATEPROGRAMWITHBINARY __clewCreateProgramWithBinary = NULL;
PFNCLCREATEPROGRAMWITHBUILTINKERNELS __clewCreateProgramWithBuiltInKernels = NULL;
PFNCLRETAINPROGRAM __clewRetainProgram = NULL;
PFNCLRELEASEPROGRAM __clewReleaseProgram = NULL;
PFNCLBUILDPROGRAM __clewBuildProgram = NULL;
PFNCLGETPROGRAMINFO __clewGetProgramInfo = NULL;
PFNCLGETPROGRAMBUILDINFO __clewGetProgramBuildInfo = NULL;
PFNCLCREATEKERNEL __clewCreateKernel = NULL;
PFNCLCREATEKERNELSINPROGRAM __clewCreateKernelsInProgram = NULL;
PFNCLRETAINKERNEL __clewRetainKernel = NULL;
PFNCLRELEASEKERNEL __clewReleaseKernel = NULL;
PFNCLSETKERNELARG __clewSetKernelArg = NULL;
PFNCLGETKERNELINFO __clewGetKernelInfo = NULL;
PFNCLGETKERNELWORKGROUPINFO __clewGetKernelWorkGroupInfo = NULL;
PFNCLWAITFOREVENTS __clewWaitForEvents = NULL;
PFNCLGETEVENTINFO __clewGetEventInfo = NULL;
PFNCLCREATEUSEREVENT __clewCreateUserEvent = NULL;
PFNCLRETAINEVENT __clewRetainEvent = NULL;
PFNCLRELEASEEVENT __clewReleaseEvent = NULL;
PFNCLSETUSEREVENTSTATUS __clewSetUserEventStatus = NULL;
PFNCLSETEVENTCALLBACK __clewSetEventCallback = NULL;
PFNCLGETEVENTPROFILINGINFO __clewGetEventProfilingInfo = NULL;
PFNCLFLUSH __clewFlush = NULL;
PFNCLFINISH __clewFinish = NULL;
PFNCLENQUEUEREADBUFFER __clewEnqueueReadBuffer = NULL;
PFNCLENQUEUEREADBUFFERRECT __clewEnqueueReadBufferRect = NULL;
PFNCLENQUEUEWRITEBUFFER __clewEnqueueWriteBuffer = NULL;
PFNCLENQUEUEWRITEBUFFERRECT __clewEnqueueWriteBufferRect = NULL;
PFNCLENQUEUECOPYBUFFER __clewEnqueueCopyBuffer = NULL;
PFNCLENQUEUEREADIMAGE __clewEnqueueReadImage = NULL;
PFNCLENQUEUEWRITEIMAGE __clewEnqueueWriteImage = NULL;
PFNCLENQUEUECOPYIMAGE __clewEnqueueCopyImage = NULL;
PFNCLENQUEUECOPYBUFFERRECT __clewEnqueueCopyBufferRect = NULL;
PFNCLENQUEUECOPYIMAGETOBUFFER __clewEnqueueCopyImageToBuffer = NULL;
PFNCLENQUEUECOPYBUFFERTOIMAGE __clewEnqueueCopyBufferToImage = NULL;
PFNCLENQUEUEMAPBUFFER __clewEnqueueMapBuffer = NULL;
PFNCLENQUEUEMAPIMAGE __clewEnqueueMapImage = NULL;
PFNCLENQUEUEUNMAPMEMOBJECT __clewEnqueueUnmapMemObject = NULL;
PFNCLENQUEUENDRANGEKERNEL __clewEnqueueNDRangeKernel = NULL;
PFNCLENQUEUETASK __clewEnqueueTask = NULL;
PFNCLENQUEUENATIVEKERNEL __clewEnqueueNativeKernel = NULL;
PFNCLGETEXTENSIONFUNCTIONADDRESSFORPLATFORM __clewGetExtensionFunctionAddressForPlatform = NULL;
#ifdef CL_USE_DEPRECATED_OPENCL_1_1_APIS
PFNCLCREATEIMAGE2D __clewCreateImage2D = NULL;
PFNCLCREATEIMAGE3D __clewCreateImage3D = NULL;
PFNCLENQUEUEMARKER __clewEnqueueMarker = NULL;
PFNCLENQUEUEWAITFOREVENTS __clewEnqueueWaitForEvents = NULL;
PFNCLENQUEUEBARRIER __clewEnqueueBarrier = NULL;
PFNCLUNLOADCOMPILER __clewUnloadCompiler = NULL;
PFNCLGETEXTENSIONFUNCTIONADDRESS __clewGetExtensionFunctionAddress = NULL;
#endif
/* cl_gl */
PFNCLCREATEFROMGLBUFFER __clewCreateFromGLBuffer = NULL;
PFNCLCREATEFROMGLTEXTURE __clewCreateFromGLTexture = NULL;
PFNCLCREATEFROMGLRENDERBUFFER __clewCreateFromGLRenderbuffer = NULL;
PFNCLGETGLOBJECTINFO __clewGetGLObjectInfo = NULL;
PFNCLGETGLTEXTUREINFO __clewGetGLTextureInfo = NULL;
PFNCLENQUEUEACQUIREGLOBJECTS __clewEnqueueAcquireGLObjects = NULL;
PFNCLENQUEUERELEASEGLOBJECTS __clewEnqueueReleaseGLObjects = NULL;
#ifdef CL_USE_DEPRECATED_OPENCL_1_1_APIS
PFNCLCREATEFROMGLTEXTURE2D __clewCreateFromGLTexture2D = NULL;
PFNCLCREATEFROMGLTEXTURE3D __clewCreateFromGLTexture3D = NULL;
#endif
PFNCLGETGLCONTEXTINFOKHR __clewGetGLContextInfoKHR = NULL;
static CLEW_DYNLIB_HANDLE dynamic_library_open_find(const char **paths) {
int i = 0;
while (paths[i] != NULL) {
CLEW_DYNLIB_HANDLE lib = CLEW_DYNLIB_OPEN(paths[i]);
if (lib != NULL) {
return lib;
}
++i;
}
return NULL;
}
static void clewExit(void)
{
if (module != NULL)
{
// Ignore errors
CLEW_DYNLIB_CLOSE(module);
module = NULL;
}
}
int clewInit()
{
#ifdef _WIN32
const char *paths[] = {"OpenCL.dll", NULL};
#elif defined(__APPLE__)
const char *paths[] = {"/Library/Frameworks/OpenCL.framework/OpenCL", NULL};
#else
const char *paths[] = {"libOpenCL.so",
"libOpenCL.so.0",
"libOpenCL.so.1",
"libOpenCL.so.2",
NULL};
#endif
int error = 0;
// Check if already initialized
if (module != NULL)
{
return CLEW_SUCCESS;
}
// Load library
module = dynamic_library_open_find(paths);
// Check for errors
if (module == NULL)
{
return CLEW_ERROR_OPEN_FAILED;
}
// Set unloading
error = atexit(clewExit);
if (error)
{
// Failure queuing atexit, shutdown with error
CLEW_DYNLIB_CLOSE(module);
module = NULL;
return CLEW_ERROR_ATEXIT_FAILED;
}
// Determine function entry-points
__clewGetPlatformIDs = (PFNCLGETPLATFORMIDS )CLEW_DYNLIB_IMPORT(module, "clGetPlatformIDs");
__clewGetPlatformInfo = (PFNCLGETPLATFORMINFO )CLEW_DYNLIB_IMPORT(module, "clGetPlatformInfo");
__clewGetDeviceIDs = (PFNCLGETDEVICEIDS )CLEW_DYNLIB_IMPORT(module, "clGetDeviceIDs");
__clewGetDeviceInfo = (PFNCLGETDEVICEINFO )CLEW_DYNLIB_IMPORT(module, "clGetDeviceInfo");
__clewCreateSubDevices = (PFNCLCREATESUBDEVICES )CLEW_DYNLIB_IMPORT(module, "clCreateSubDevices");
__clewRetainDevice = (PFNCLRETAINDEVICE )CLEW_DYNLIB_IMPORT(module, "clRetainDevice");
__clewReleaseDevice = (PFNCLRELEASEDEVICE )CLEW_DYNLIB_IMPORT(module, "clReleaseDevice");
__clewCreateContext = (PFNCLCREATECONTEXT )CLEW_DYNLIB_IMPORT(module, "clCreateContext");
__clewCreateContextFromType = (PFNCLCREATECONTEXTFROMTYPE )CLEW_DYNLIB_IMPORT(module, "clCreateContextFromType");
__clewRetainContext = (PFNCLRETAINCONTEXT )CLEW_DYNLIB_IMPORT(module, "clRetainContext");
__clewReleaseContext = (PFNCLRELEASECONTEXT )CLEW_DYNLIB_IMPORT(module, "clReleaseContext");
__clewGetContextInfo = (PFNCLGETCONTEXTINFO )CLEW_DYNLIB_IMPORT(module, "clGetContextInfo");
__clewCreateCommandQueue = (PFNCLCREATECOMMANDQUEUE )CLEW_DYNLIB_IMPORT(module, "clCreateCommandQueue");
__clewRetainCommandQueue = (PFNCLRETAINCOMMANDQUEUE )CLEW_DYNLIB_IMPORT(module, "clRetainCommandQueue");
__clewReleaseCommandQueue = (PFNCLRELEASECOMMANDQUEUE )CLEW_DYNLIB_IMPORT(module, "clReleaseCommandQueue");
__clewGetCommandQueueInfo = (PFNCLGETCOMMANDQUEUEINFO )CLEW_DYNLIB_IMPORT(module, "clGetCommandQueueInfo");
#ifdef CL_USE_DEPRECATED_OPENCL_1_0_APIS
__clewSetCommandQueueProperty = (PFNCLSETCOMMANDQUEUEPROPERTY )CLEW_DYNLIB_IMPORT(module, "clSetCommandQueueProperty");
#endif
__clewCreateBuffer = (PFNCLCREATEBUFFER )CLEW_DYNLIB_IMPORT(module, "clCreateBuffer");
__clewCreateSubBuffer = (PFNCLCREATESUBBUFFER )CLEW_DYNLIB_IMPORT(module, "clCreateSubBuffer");
__clewCreateImage = (PFNCLCREATEIMAGE )CLEW_DYNLIB_IMPORT(module, "clCreateImage");
__clewRetainMemObject = (PFNCLRETAINMEMOBJECT )CLEW_DYNLIB_IMPORT(module, "clRetainMemObject");
__clewReleaseMemObject = (PFNCLRELEASEMEMOBJECT )CLEW_DYNLIB_IMPORT(module, "clReleaseMemObject");
__clewGetSupportedImageFormats = (PFNCLGETSUPPORTEDIMAGEFORMATS )CLEW_DYNLIB_IMPORT(module, "clGetSupportedImageFormats");
__clewGetMemObjectInfo = (PFNCLGETMEMOBJECTINFO )CLEW_DYNLIB_IMPORT(module, "clGetMemObjectInfo");
__clewGetImageInfo = (PFNCLGETIMAGEINFO )CLEW_DYNLIB_IMPORT(module, "clGetImageInfo");
__clewSetMemObjectDestructorCallback = (PFNCLSETMEMOBJECTDESTRUCTORCALLBACK)CLEW_DYNLIB_IMPORT(module, "clSetMemObjectDestructorCallback");
__clewCreateSampler = (PFNCLCREATESAMPLER )CLEW_DYNLIB_IMPORT(module, "clCreateSampler");
__clewRetainSampler = (PFNCLRETAINSAMPLER )CLEW_DYNLIB_IMPORT(module, "clRetainSampler");
__clewReleaseSampler = (PFNCLRELEASESAMPLER )CLEW_DYNLIB_IMPORT(module, "clReleaseSampler");
__clewGetSamplerInfo = (PFNCLGETSAMPLERINFO )CLEW_DYNLIB_IMPORT(module, "clGetSamplerInfo");
__clewCreateProgramWithSource = (PFNCLCREATEPROGRAMWITHSOURCE )CLEW_DYNLIB_IMPORT(module, "clCreateProgramWithSource");
__clewCreateProgramWithBinary = (PFNCLCREATEPROGRAMWITHBINARY )CLEW_DYNLIB_IMPORT(module, "clCreateProgramWithBinary");
__clewCreateProgramWithBuiltInKernels =(PFNCLCREATEPROGRAMWITHBUILTINKERNELS)CLEW_DYNLIB_IMPORT(module, "clCreateProgramWithBuiltInKernels");
__clewRetainProgram = (PFNCLRETAINPROGRAM )CLEW_DYNLIB_IMPORT(module, "clRetainProgram");
__clewReleaseProgram = (PFNCLRELEASEPROGRAM )CLEW_DYNLIB_IMPORT(module, "clReleaseProgram");
__clewBuildProgram = (PFNCLBUILDPROGRAM )CLEW_DYNLIB_IMPORT(module, "clBuildProgram");
__clewGetProgramInfo = (PFNCLGETPROGRAMINFO )CLEW_DYNLIB_IMPORT(module, "clGetProgramInfo");
__clewGetProgramBuildInfo = (PFNCLGETPROGRAMBUILDINFO )CLEW_DYNLIB_IMPORT(module, "clGetProgramBuildInfo");
__clewCreateKernel = (PFNCLCREATEKERNEL )CLEW_DYNLIB_IMPORT(module, "clCreateKernel");
__clewCreateKernelsInProgram = (PFNCLCREATEKERNELSINPROGRAM )CLEW_DYNLIB_IMPORT(module, "clCreateKernelsInProgram");
__clewRetainKernel = (PFNCLRETAINKERNEL )CLEW_DYNLIB_IMPORT(module, "clRetainKernel");
__clewReleaseKernel = (PFNCLRELEASEKERNEL )CLEW_DYNLIB_IMPORT(module, "clReleaseKernel");
__clewSetKernelArg = (PFNCLSETKERNELARG )CLEW_DYNLIB_IMPORT(module, "clSetKernelArg");
__clewGetKernelInfo = (PFNCLGETKERNELINFO )CLEW_DYNLIB_IMPORT(module, "clGetKernelInfo");
__clewGetKernelWorkGroupInfo = (PFNCLGETKERNELWORKGROUPINFO )CLEW_DYNLIB_IMPORT(module, "clGetKernelWorkGroupInfo");
__clewWaitForEvents = (PFNCLWAITFOREVENTS )CLEW_DYNLIB_IMPORT(module, "clWaitForEvents");
__clewGetEventInfo = (PFNCLGETEVENTINFO )CLEW_DYNLIB_IMPORT(module, "clGetEventInfo");
__clewCreateUserEvent = (PFNCLCREATEUSEREVENT )CLEW_DYNLIB_IMPORT(module, "clCreateUserEvent");
__clewRetainEvent = (PFNCLRETAINEVENT )CLEW_DYNLIB_IMPORT(module, "clRetainEvent");
__clewReleaseEvent = (PFNCLRELEASEEVENT )CLEW_DYNLIB_IMPORT(module, "clReleaseEvent");
__clewSetUserEventStatus = (PFNCLSETUSEREVENTSTATUS )CLEW_DYNLIB_IMPORT(module, "clSetUserEventStatus");
__clewSetEventCallback = (PFNCLSETEVENTCALLBACK )CLEW_DYNLIB_IMPORT(module, "clSetEventCallback");
__clewGetEventProfilingInfo = (PFNCLGETEVENTPROFILINGINFO )CLEW_DYNLIB_IMPORT(module, "clGetEventProfilingInfo");
__clewFlush = (PFNCLFLUSH )CLEW_DYNLIB_IMPORT(module, "clFlush");
__clewFinish = (PFNCLFINISH )CLEW_DYNLIB_IMPORT(module, "clFinish");
__clewEnqueueReadBuffer = (PFNCLENQUEUEREADBUFFER )CLEW_DYNLIB_IMPORT(module, "clEnqueueReadBuffer");
__clewEnqueueReadBufferRect = (PFNCLENQUEUEREADBUFFERRECT )CLEW_DYNLIB_IMPORT(module, "clEnqueueReadBufferRect");
__clewEnqueueWriteBuffer = (PFNCLENQUEUEWRITEBUFFER )CLEW_DYNLIB_IMPORT(module, "clEnqueueWriteBuffer");
__clewEnqueueWriteBufferRect = (PFNCLENQUEUEWRITEBUFFERRECT )CLEW_DYNLIB_IMPORT(module, "clEnqueueWriteBufferRect");
__clewEnqueueCopyBuffer = (PFNCLENQUEUECOPYBUFFER )CLEW_DYNLIB_IMPORT(module, "clEnqueueCopyBuffer");
__clewEnqueueCopyBufferRect = (PFNCLENQUEUECOPYBUFFERRECT )CLEW_DYNLIB_IMPORT(module, "clEnqueueCopyBufferRect");
__clewEnqueueReadImage = (PFNCLENQUEUEREADIMAGE )CLEW_DYNLIB_IMPORT(module, "clEnqueueReadImage");
__clewEnqueueWriteImage = (PFNCLENQUEUEWRITEIMAGE )CLEW_DYNLIB_IMPORT(module, "clEnqueueWriteImage");
__clewEnqueueCopyImage = (PFNCLENQUEUECOPYIMAGE )CLEW_DYNLIB_IMPORT(module, "clEnqueueCopyImage");
__clewEnqueueCopyImageToBuffer = (PFNCLENQUEUECOPYIMAGETOBUFFER )CLEW_DYNLIB_IMPORT(module, "clEnqueueCopyImageToBuffer");
__clewEnqueueCopyBufferToImage = (PFNCLENQUEUECOPYBUFFERTOIMAGE )CLEW_DYNLIB_IMPORT(module, "clEnqueueCopyBufferToImage");
__clewEnqueueMapBuffer = (PFNCLENQUEUEMAPBUFFER )CLEW_DYNLIB_IMPORT(module, "clEnqueueMapBuffer");
__clewEnqueueMapImage = (PFNCLENQUEUEMAPIMAGE )CLEW_DYNLIB_IMPORT(module, "clEnqueueMapImage");
__clewEnqueueUnmapMemObject = (PFNCLENQUEUEUNMAPMEMOBJECT )CLEW_DYNLIB_IMPORT(module, "clEnqueueUnmapMemObject");
__clewEnqueueNDRangeKernel = (PFNCLENQUEUENDRANGEKERNEL )CLEW_DYNLIB_IMPORT(module, "clEnqueueNDRangeKernel");
__clewEnqueueTask = (PFNCLENQUEUETASK )CLEW_DYNLIB_IMPORT(module, "clEnqueueTask");
__clewEnqueueNativeKernel = (PFNCLENQUEUENATIVEKERNEL )CLEW_DYNLIB_IMPORT(module, "clEnqueueNativeKernel");
__clewGetExtensionFunctionAddressForPlatform = (PFNCLGETEXTENSIONFUNCTIONADDRESSFORPLATFORM)CLEW_DYNLIB_IMPORT(module, "clGetExtensionFunctionAddressForPlatform");
#ifdef CL_USE_DEPRECATED_OPENCL_1_1_APIS
__clewCreateImage2D = (PFNCLCREATEIMAGE2D )CLEW_DYNLIB_IMPORT(module, "clCreateImage2D");
__clewCreateImage3D = (PFNCLCREATEIMAGE3D )CLEW_DYNLIB_IMPORT(module, "clCreateImage3D");
__clewEnqueueMarker = (PFNCLENQUEUEMARKER )CLEW_DYNLIB_IMPORT(module, "clEnqueueMarker");
__clewEnqueueWaitForEvents = (PFNCLENQUEUEWAITFOREVENTS )CLEW_DYNLIB_IMPORT(module, "clEnqueueWaitForEvents");
__clewEnqueueBarrier = (PFNCLENQUEUEBARRIER )CLEW_DYNLIB_IMPORT(module, "clEnqueueBarrier");
__clewUnloadCompiler = (PFNCLUNLOADCOMPILER )CLEW_DYNLIB_IMPORT(module, "clUnloadCompiler");
__clewGetExtensionFunctionAddress = (PFNCLGETEXTENSIONFUNCTIONADDRESS )CLEW_DYNLIB_IMPORT(module, "clGetExtensionFunctionAddress");
#endif
/* cl_gl */
__clewCreateFromGLBuffer = (PFNCLCREATEFROMGLBUFFER )CLEW_DYNLIB_IMPORT(module, "clCreateFromGLBuffer");
__clewCreateFromGLTexture = (PFNCLCREATEFROMGLTEXTURE )CLEW_DYNLIB_IMPORT(module, "clCreateFromGLTexture");
__clewCreateFromGLRenderbuffer = (PFNCLCREATEFROMGLRENDERBUFFER )CLEW_DYNLIB_IMPORT(module, "clCreateFromGLRenderbuffer");
__clewGetGLObjectInfo = (PFNCLGETGLOBJECTINFO )CLEW_DYNLIB_IMPORT(module, "clGetGLObjectInfo");
__clewGetGLTextureInfo = (PFNCLGETGLTEXTUREINFO )CLEW_DYNLIB_IMPORT(module, "clGetGLTextureInfo");
__clewEnqueueAcquireGLObjects = (PFNCLENQUEUEACQUIREGLOBJECTS )CLEW_DYNLIB_IMPORT(module, "clEnqueueAcquireGLObjects");
__clewEnqueueReleaseGLObjects = (PFNCLENQUEUERELEASEGLOBJECTS )CLEW_DYNLIB_IMPORT(module, "clEnqueueReleaseGLObjects");
#ifdef CL_USE_DEPRECATED_OPENCL_1_1_APIS
__clewCreateFromGLTexture2D = (PFNCLCREATEFROMGLTEXTURE2D )CLEW_DYNLIB_IMPORT(module, "clCreateFromGLTexture2D");
__clewCreateFromGLTexture3D = (PFNCLCREATEFROMGLTEXTURE3D )CLEW_DYNLIB_IMPORT(module, "clCreateFromGLTexture3D");
#endif
__clewGetGLContextInfoKHR = (PFNCLGETGLCONTEXTINFOKHR )CLEW_DYNLIB_IMPORT(module, "clGetGLContextInfoKHR");
if(__clewGetPlatformIDs == NULL) return 0;
if(__clewGetPlatformInfo == NULL) return 0;
if(__clewGetDeviceIDs == NULL) return 0;
if(__clewGetDeviceInfo == NULL) return 0;
return CLEW_SUCCESS;
}
const char* clewErrorString(cl_int error)
{
static const char* strings[] =
{
// Error Codes
"CL_SUCCESS" // 0
, "CL_DEVICE_NOT_FOUND" // -1
, "CL_DEVICE_NOT_AVAILABLE" // -2
, "CL_COMPILER_NOT_AVAILABLE" // -3
, "CL_MEM_OBJECT_ALLOCATION_FAILURE" // -4
, "CL_OUT_OF_RESOURCES" // -5
, "CL_OUT_OF_HOST_MEMORY" // -6
, "CL_PROFILING_INFO_NOT_AVAILABLE" // -7
, "CL_MEM_COPY_OVERLAP" // -8
, "CL_IMAGE_FORMAT_MISMATCH" // -9
, "CL_IMAGE_FORMAT_NOT_SUPPORTED" // -10
, "CL_BUILD_PROGRAM_FAILURE" // -11
, "CL_MAP_FAILURE" // -12
, "CL_MISALIGNED_SUB_BUFFER_OFFSET" // -13
, "CL_EXEC_STATUS_ERROR_FOR_EVENTS_IN_WAIT_LIST"// -14
, "CL_COMPILE_PROGRAM_FAILURE" // -15
, "CL_LINKER_NOT_AVAILABLE" // -16
, "CL_LINK_PROGRAM_FAILURE" // -17
, "CL_DEVICE_PARTITION_FAILED" // -18
, "CL_KERNEL_ARG_INFO_NOT_AVAILABLE" // -19
, "" // -20
, "" // -21
, "" // -22
, "" // -23
, "" // -24
, "" // -25
, "" // -26
, "" // -27
, "" // -28
, "" // -29
, "CL_INVALID_VALUE" // -30
, "CL_INVALID_DEVICE_TYPE" // -31
, "CL_INVALID_PLATFORM" // -32
, "CL_INVALID_DEVICE" // -33
, "CL_INVALID_CONTEXT" // -34
, "CL_INVALID_QUEUE_PROPERTIES" // -35
, "CL_INVALID_COMMAND_QUEUE" // -36
, "CL_INVALID_HOST_PTR" // -37
, "CL_INVALID_MEM_OBJECT" // -38
, "CL_INVALID_IMAGE_FORMAT_DESCRIPTOR" // -39
, "CL_INVALID_IMAGE_SIZE" // -40
, "CL_INVALID_SAMPLER" // -41
, "CL_INVALID_BINARY" // -42
, "CL_INVALID_BUILD_OPTIONS" // -43
, "CL_INVALID_PROGRAM" // -44
, "CL_INVALID_PROGRAM_EXECUTABLE" // -45
, "CL_INVALID_KERNEL_NAME" // -46
, "CL_INVALID_KERNEL_DEFINITION" // -47
, "CL_INVALID_KERNEL" // -48
, "CL_INVALID_ARG_INDEX" // -49
, "CL_INVALID_ARG_VALUE" // -50
, "CL_INVALID_ARG_SIZE" // -51
, "CL_INVALID_KERNEL_ARGS" // -52
, "CL_INVALID_WORK_DIMENSION" // -53
, "CL_INVALID_WORK_GROUP_SIZE" // -54
, "CL_INVALID_WORK_ITEM_SIZE" // -55
, "CL_INVALID_GLOBAL_OFFSET" // -56
, "CL_INVALID_EVENT_WAIT_LIST" // -57
, "CL_INVALID_EVENT" // -58
, "CL_INVALID_OPERATION" // -59
, "CL_INVALID_GL_OBJECT" // -60
, "CL_INVALID_BUFFER_SIZE" // -61
, "CL_INVALID_MIP_LEVEL" // -62
, "CL_INVALID_GLOBAL_WORK_SIZE" // -63
, "CL_INVALID_PROPERTY" // -64
, "CL_INVALID_IMAGE_DESCRIPTOR" // -65
, "CL_INVALID_COMPILER_OPTIONS" // -66
, "CL_INVALID_LINKER_OPTIONS" // -67
, "CL_INVALID_DEVICE_PARTITION_COUNT" // -68
};
static const int num_errors = sizeof(strings) / sizeof(strings[0]);
if (error == -1001) {
return "CL_PLATFORM_NOT_FOUND_KHR";
}
if (error > 0 || -error >= num_errors) {
return "Unknown OpenCL error";
}
return strings[-error];
}

1
release/license/THIRD-PARTY-LICENSES.txt
View File

@ -32,7 +32,6 @@ https://github.com/AcademySoftwareFoundation/MaterialX
** meson; version 0.63 -- https://github.com/mesonbuild/meson
** oneAPI Threading Building Block; version 2020_U3 --
https://software.intel.com/en-us/oneapi/onetbb
** OpenCL Wrangler; version 27a6867 -- https://github.com/OpenCLWrangler/clew
** OpenImageDenoise; version 1.4.3 -- https://www.openimagedenoise.org/
** OpenImageIO; version 2.4.15.0 -- http://www.openimageio.org
** OpenSSL; version 3.1.2 -- https://www.openssl.org/

42
source/blender/compositor/CMakeLists.txt
View File

@ -22,7 +22,6 @@ if(WITH_COMPOSITOR_CPU)
../nodes/intern
../render
../render/intern
../../../extern/clew/include
# RNA_prototypes.h
${CMAKE_BINARY_DIR}/source/blender/makesrna
@ -38,16 +37,10 @@ if(WITH_COMPOSITOR_CPU)
COM_profile.hh
intern/COM_BufferArea.h
intern/COM_BufferOperation.cc
intern/COM_BufferOperation.h
intern/COM_BufferRange.h
intern/COM_BuffersIterator.h
intern/COM_CPUDevice.cc
intern/COM_CPUDevice.h
intern/COM_ChunkOrder.cc
intern/COM_ChunkOrder.h
intern/COM_ChunkOrderHotspot.cc
intern/COM_ChunkOrderHotspot.h
intern/COM_CompositorContext.cc
intern/COM_CompositorContext.h
intern/COM_ConstantFolder.cc
@ -60,8 +53,6 @@ if(WITH_COMPOSITOR_CPU)
intern/COM_Device.h
intern/COM_Enums.cc
intern/COM_Enums.h
intern/COM_ExecutionGroup.cc
intern/COM_ExecutionGroup.h
intern/COM_ExecutionModel.cc
intern/COM_ExecutionModel.h
intern/COM_ExecutionSystem.cc
@ -70,8 +61,6 @@ if(WITH_COMPOSITOR_CPU)
intern/COM_FullFrameExecutionModel.h
intern/COM_MemoryBuffer.cc
intern/COM_MemoryBuffer.h
intern/COM_MemoryProxy.cc
intern/COM_MemoryProxy.h
intern/COM_MetaData.cc
intern/COM_MetaData.h
intern/COM_MultiThreadedOperation.cc
@ -88,15 +77,8 @@ if(WITH_COMPOSITOR_CPU)
intern/COM_NodeOperation.h
intern/COM_NodeOperationBuilder.cc
intern/COM_NodeOperationBuilder.h
intern/COM_OpenCLDevice.cc
intern/COM_OpenCLDevice.h
intern/COM_SharedOperationBuffers.cc
intern/COM_SharedOperationBuffers.h
intern/COM_SingleThreadedOperation.cc
intern/COM_SingleThreadedOperation.h
intern/COM_TiledExecutionModel.cc
intern/COM_TiledExecutionModel.h
intern/COM_WorkPackage.cc
intern/COM_WorkPackage.h
intern/COM_WorkScheduler.cc
intern/COM_WorkScheduler.h
@ -335,18 +317,10 @@ if(WITH_COMPOSITOR_CPU)
operations/COM_GammaCorrectOperation.h
operations/COM_GaussianAlphaBlurBaseOperation.cc
operations/COM_GaussianAlphaBlurBaseOperation.h
operations/COM_GaussianAlphaXBlurOperation.cc
operations/COM_GaussianAlphaXBlurOperation.h
operations/COM_GaussianAlphaYBlurOperation.cc
operations/COM_GaussianAlphaYBlurOperation.h
operations/COM_GaussianBlurBaseOperation.cc
operations/COM_GaussianBlurBaseOperation.h
operations/COM_GaussianBokehBlurOperation.cc
operations/COM_GaussianBokehBlurOperation.h
operations/COM_GaussianXBlurOperation.cc
operations/COM_GaussianXBlurOperation.h
operations/COM_GaussianYBlurOperation.cc
operations/COM_GaussianYBlurOperation.h
operations/COM_KuwaharaAnisotropicOperation.cc
operations/COM_KuwaharaAnisotropicOperation.h
operations/COM_KuwaharaAnisotropicStructureTensorOperation.cc
@ -473,16 +447,12 @@ if(WITH_COMPOSITOR_CPU)
operations/COM_GammaOperation.h
operations/COM_MixOperation.cc
operations/COM_MixOperation.h
operations/COM_ReadBufferOperation.cc
operations/COM_ReadBufferOperation.h
operations/COM_SetColorOperation.cc
operations/COM_SetColorOperation.h
operations/COM_SetValueOperation.cc
operations/COM_SetValueOperation.h
operations/COM_SetVectorOperation.cc
operations/COM_SetVectorOperation.h
operations/COM_WriteBufferOperation.cc
operations/COM_WriteBufferOperation.h
operations/COM_MathBaseOperation.cc
operations/COM_MathBaseOperation.h
@ -536,8 +506,6 @@ if(WITH_COMPOSITOR_CPU)
operations/COM_TransformOperation.h
operations/COM_TranslateOperation.cc
operations/COM_TranslateOperation.h
operations/COM_WrapOperation.cc
operations/COM_WrapOperation.h
# Filter operations
operations/COM_ConvolutionEdgeFilterOperation.cc
@ -607,7 +575,6 @@ if(WITH_COMPOSITOR_CPU)
PRIVATE bf::intern::clog
PRIVATE bf::intern::guardedalloc
bf_realtime_compositor
extern_clew
PRIVATE bf::intern::atomic
)
@ -615,15 +582,6 @@ if(WITH_COMPOSITOR_CPU)
${CMAKE_CURRENT_BINARY_DIR}/operations
)
data_to_c(
${CMAKE_CURRENT_SOURCE_DIR}/operations/COM_OpenCLKernels.cl
${CMAKE_CURRENT_BINARY_DIR}/operations/COM_OpenCLKernels.cl.h
SRC
STRIP_LEADING_C_COMMENTS
)
add_definitions(-DCL_USE_DEPRECATED_OPENCL_1_1_APIS)
set(GENSRC_DIR ${CMAKE_CURRENT_BINARY_DIR}/operations)
set(GENSRC ${GENSRC_DIR}/COM_SMAAAreaTexture.h)
add_custom_command(

230
source/blender/compositor/COM_compositor.hh
View File

@ -33,49 +33,6 @@ struct Render;
* \defgroup Operation All operations of the compositor
* \ingroup compositor
*
* \page Introduction of the Blender Compositor
*
* \section bcomp Blender compositor
* This project redesigns the internals of Blender's compositor.
* The project has been executed in 2011 by At Mind.
* At Mind is a technology company located in Amsterdam, The Netherlands.
* The project has been crowd-funded. This code has been released under GPL2 to be used in Blender.
*
* \section goals The goals of the project
* the new compositor has 2 goals.
* - Make a faster compositor (speed of calculation)
* - Make the compositor work faster for you (workflow)
*
* \section speed Faster compositor
* The speedup has been done by making better use of the hardware Blenders is working on.
* The previous compositor only used a single threaded model to calculate a node.
* The only exception to this is the Defocus node.
* Only when it is possible to calculate two full nodes in parallel a second thread was used.
* Current workstations have 8-16 threads available, and most of the time these are idle.
*
* In the new compositor we want to use as much of threads as possible.
* Even new OpenCL capable GPU-hardware can be used for calculation.
*
* \section workflow Work faster
* The previous compositor only showed the final image.
* The compositor could wait a long time before seeing the result of his work.
* The new compositor will work in a way that it will focus on
* getting information back to the user. It will prioritize its work to get earlier user feedback.
*
* \page memory Memory model
* The main issue is the type of memory model to use.
* Blender is used by consumers and professionals.
* Ranging from low-end machines to very high-end machines.
* The system should work on high-end machines and on low-end machines.
* \page executing Executing
* \section prepare Prepare execution
*
* during the preparation of the execution All ReadBufferOperation will receive an offset.
* This offset is used during execution as an optimization trick
* Next all operations will be initialized for execution \see NodeOperation.init_execution
* Next all ExecutionGroup's will be initialized for execution \see ExecutionGroup.init_execution
* this all is controlled from \see ExecutionSystem.execute
*
* \section priority Render priority
* Render priority is an priority of an output node.
* A user has a different need of Render priorities of output nodes
@ -85,152 +42,6 @@ struct Render;
* All NodeOperation has a setting for their render-priority,
* but only for output NodeOperation these have effect.
* In ExecutionSystem.execute all priorities are checked.
* For every priority the ExecutionGroup's are check if the
* priority do match.
* When match the ExecutionGroup will be executed (this happens in serial)
*
* \see ExecutionSystem.execute control of the Render priority
* \see NodeOperation.get_render_priority receive the render priority
* \see ExecutionGroup.execute the main loop to execute a whole ExecutionGroup
*
* \section order Chunk order
*
* When a ExecutionGroup is executed, first the order of chunks are determined.
* The settings are stored in the ViewerNode inside the ExecutionGroup.
* ExecutionGroups that have no viewer-node,
* will use a default one.
* There are several possible chunk orders
* - [@ref ChunkOrdering.CenterOut]:
* Start calculating from a configurable point and order by nearest chunk.
* - [@ref ChunkOrdering.Random]:
* Randomize all chunks.
* - [@ref ChunkOrdering.TopDown]:
* Start calculation from the bottom to the top of the image.
* - [@ref ChunkOrdering.RuleOfThirds]:
* Experimental order based on 9 hot-spots in the image.
*
* When the chunk-order is determined, the first few chunks will be checked if they can be scheduled.
* Chunks can have three states:
* - [@ref eWorkPackageState.NotScheduled]:
* Chunk is not yet scheduled, or dependencies are not met.
* - [@ref eWorkPackageState.Scheduled]:
* All dependencies are met, chunk is scheduled, but not finished.
* - [@ref eWorkPackageState.Executed]:
* Chunk is finished.
*
* \see ExecutionGroup.execute
* \see ViewerOperation.get_chunk_order
* \see ChunkOrdering
*
* \section interest Area of interest
* An ExecutionGroup can have dependencies to other ExecutionGroup's.
* Data passing from one ExecutionGroup to another one are stored in 'chunks'.
* If not all input chunks are available the chunk execution will not be scheduled.
* <pre>
* +-------------------------------------+ +--------------------------------------+
* | ExecutionGroup A | | ExecutionGroup B |
* | +----------------+ +-------------+ | | +------------+ +-----------------+ |
* | | NodeOperation a| | WriteBuffer | | | | ReadBuffer | | ViewerOperation | |
* | | *==* Operation | | | | Operation *===* | |
* | | | | | | | | | | | |
* | +----------------+ +-------------+ | | +------------+ +-----------------+ |
* | | | | | |
* +--------------------------------|----+ +---|----------------------------------+
* | |
* | |
* +---------------------------+
* | MemoryProxy |
* | +----------+ +---------+ |
* | | Chunk a | | Chunk b | |
* | | | | | |
* | +----------+ +---------+ |
* | |
* +---------------------------+
* </pre>
*
* In the above example ExecutionGroup B has an outputoperation (ViewerOperation)
* and is being executed.
* The first chunk is evaluated [@ref ExecutionGroup.schedule_chunk_when_possible],
* but not all input chunks are available.
* The relevant ExecutionGroup (that can calculate the missing chunks; ExecutionGroup A)
* is asked to calculate the area ExecutionGroup B is missing.
* [@ref ExecutionGroup.schedule_area_when_possible]
* ExecutionGroup B checks what chunks the area spans, and tries to schedule these chunks.
* If all input data is available these chunks are scheduled [@ref ExecutionGroup.schedule_chunk]
*
* <pre>
*
* +-------------------------+ +----------------+ +----------------+
* | ExecutionSystem.execute | | ExecutionGroup | | ExecutionGroup |
* +-------------------------+ | (B) | | (A) |
* O +----------------+ +----------------+
* O | |
* O ExecutionGroup.execute | |
* O------------------------------->O |
* . O |
* . O-------\ |
* . . | ExecutionGroup.schedule_chunk_when_possible
* . . O----/ (*) |
* . . O |
* . . O |
* . . O ExecutionGroup.schedule_area_when_possible|
* . . O---------------------------------------->O
* . . . O----------\ ExecutionGroup.schedule_chunk_when_possible
* . . . . | (*)
* . . . . O-------/
* . . . . O
* . . . . O
* . . . . O-------\ ExecutionGroup.schedule_chunk
* . . . . . |
* . . . . . O----/
* . . . . O<=O
* . . . O<=O
* . . . O
* . . O<========================================O
* . . O |
* . O<=O |
* . O |
* . O |
* </pre>
*
* This happens until all chunks of (ExecutionGroup B) are finished executing or the user break's the process.
*
* NodeOperation like the ScaleOperation can influence the area of interest by reimplementing the
* [@ref NodeOperation.determine_area_of_interest] method
*
* <pre>
*
* +--------------------------+ +---------------------------------+
* | ExecutionGroup A | | ExecutionGroup B |
* | | | |
* +--------------------------+ +---------------------------------+
* Needed chunks from ExecutionGroup A | Chunk of ExecutionGroup B (to be evaluated)
* +-------+ +-------+ | +--------+
* |Chunk 1| |Chunk 2| +----------------+ |Chunk 1 |
* | | | | | ScaleOperation | | |
* +-------+ +-------+ +----------------+ +--------+
*
* +-------+ +-------+
* |Chunk 3| |Chunk 4|
* | | | |
* +-------+ +-------+
*
* </pre>
*
* \see ExecutionGroup.execute Execute a complete ExecutionGroup.
* Halts until finished or breaked by user
* \see ExecutionGroup.schedule_chunk_when_possible Tries to schedule a single chunk,
* checks if all input data is available. Can trigger dependent chunks to be calculated
* \see ExecutionGroup.schedule_area_when_possible
* Tries to schedule an area. This can be multiple chunks
* (is called from [@ref ExecutionGroup.schedule_chunk_when_possible])
* \see ExecutionGroup.schedule_chunk Schedule a chunk on the WorkScheduler
* \see NodeOperation.determine_depending_area_of_interest Influence the area of interest of a chunk.
* \see WriteBufferOperation Operation to write to a MemoryProxy/MemoryBuffer
* \see ReadBufferOperation Operation to read from a MemoryProxy/MemoryBuffer
* \see MemoryProxy proxy for information about memory image
* (a image consist out of multiple chunks)
* \see MemoryBuffer Allocated memory for a single chunk
*
* \section workscheduler WorkScheduler
* the WorkScheduler is implemented as a static class. the responsibility of the WorkScheduler
@ -251,47 +62,6 @@ struct Render;
* This is done by changing the `COM_threading_model`
* to `ThreadingModel::SingleThreaded`. When compiling the work-scheduler
* will be changes to support no threading and run everything on the CPU.
*
* \section devices Devices
* A Device within the compositor context is a Hardware component that can used to calculate chunks.
* This chunk is encapsulated in a WorkPackage.
* the WorkScheduler controls the devices and selects the device where a
* WorkPackage will be calculated.
*
* \subsection WS_Devices Work-scheduler
* The WorkScheduler controls all Devices.
* When initializing the compositor the WorkScheduler selects all
* devices that will be used during compositor.
* There are two types of Devices, CPUDevice and OpenCLDevice.
* When an ExecutionGroup schedules a Chunk the schedule method of the WorkScheduler
* The Workscheduler determines if the chunk can be run on an OpenCLDevice
* (and that there are available OpenCLDevice).
* If this is the case the chunk will be added to the work-list for OpenCLDevice's
* otherwise the chunk will be added to the work-list of CPUDevices.
*
* A thread will read the work-list and sends a work-package to its device.
*
* \see WorkScheduler.schedule method that is called to schedule a chunk
* \see Device.execute method called to execute a chunk
*
* \subsection CPUDevice CPUDevice
* When a CPUDevice gets a WorkPackage the Device will get the input-buffer that is needed to
* calculate the chunk. Allocation is already done by the ExecutionGroup.
* The output-buffer of the chunk is being created.
* The OutputOperation of the ExecutionGroup is called to execute the area of the output-buffer.
*
* \see ExecutionGroup
* \see NodeOperation.execute_region executes a single chunk of a NodeOperation
* \see CPUDevice.execute
*
* \subsection GPUDevice OpenCLDevice
*
* To be completed!
* \see NodeOperation.execute_opencl_region
* \see OpenCLDevice.execute
*
* \section execute_pixel executing a pixel
* Finally the last step, the node functionality :)
*/
/**

32
source/blender/compositor/COM_defines.h
View File

@ -12,16 +12,6 @@ namespace blender::compositor {
using Size2f = float2;
enum class eExecutionModel {
/**
* Operations are executed from outputs to inputs grouped in execution groups and rendered
* in tiles.
*/
Tiled,
/** Operations are fully rendered in order from inputs to outputs. */
FullFrame
};
enum class eDimension { X, Y };
/**
@ -84,28 +74,6 @@ constexpr DataType COM_num_channels_data_type(const int num_channels)
}
}
/* Configurable items.
*
* Chunk size determination.
*
* Chunk order. */
/**
* \brief The order of chunks to be scheduled
* \ingroup Execution
*/
enum class ChunkOrdering {
/** \brief order from a distance to centerX/centerY */
CenterOut = 0,
/** \brief order randomly */
Random = 1,
/** \brief no ordering */
TopDown = 2,
/** \brief experimental ordering with 9 hot-spots. */
RuleOfThirds = 3,
Default = ChunkOrdering::CenterOut,
};
constexpr float COM_PREVIEW_SIZE = 140.f;
constexpr float COM_RULE_OF_THIRDS_DIVIDER = 100.0f;
constexpr float COM_BLUR_BOKEH_PIXELS = 512;

2
source/blender/compositor/COM_precomp.h
View File

@ -23,12 +23,10 @@
#include "COM_ConvertOperation.h"
#include "COM_Debug.h"
#include "COM_Enums.h"
#include "COM_ExecutionGroup.h"
#include "COM_ExecutionSystem.h"
#include "COM_MultiThreadedOperation.h"
#include "COM_Node.h"
#include "COM_NodeOperation.h"
#include "COM_OpenCLDevice.h"
#include "COM_SetAlphaMultiplyOperation.h"
#include "COM_SetColorOperation.h"
#include "COM_SetSamplerOperation.h"

82
source/blender/compositor/intern/COM_BufferOperation.cc
View File

@ -1,82 +0,0 @@
/* SPDX-FileCopyrightText: 2021 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_BufferOperation.h"
namespace blender::compositor {
BufferOperation::BufferOperation(MemoryBuffer *buffer, DataType data_type)
{
buffer_ = buffer;
inflated_buffer_ = nullptr;
set_canvas(buffer->get_rect());
add_output_socket(data_type);
flags_.is_constant_operation = buffer_->is_a_single_elem();
flags_.is_fullframe_operation = false;
}
const float *BufferOperation::get_constant_elem()
{
BLI_assert(buffer_->is_a_single_elem());
return buffer_->get_buffer();
}
void BufferOperation::init_execution()
{
if (buffer_->is_a_single_elem()) {
init_mutex();
}
}
void *BufferOperation::initialize_tile_data(rcti * /*rect*/)
{
if (buffer_->is_a_single_elem() == false) {
return buffer_;
}
lock_mutex();
if (!inflated_buffer_) {
inflated_buffer_ = buffer_->inflate();
}
unlock_mutex();
return inflated_buffer_;
}
void BufferOperation::deinit_execution()
{
if (buffer_->is_a_single_elem()) {
deinit_mutex();
}
delete inflated_buffer_;
}
void BufferOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
switch (sampler) {
case PixelSampler::Nearest:
buffer_->read(output, x, y);
break;
case PixelSampler::Bilinear:
default:
buffer_->read_bilinear(output, x, y);
break;
case PixelSampler::Bicubic:
/* No bicubic. Same implementation as ReadBufferOperation. */
buffer_->read_bilinear(output, x, y);
break;
}
}
void BufferOperation::execute_pixel_filtered(
float output[4], float x, float y, float dx[2], float dy[2])
{
const float uv[2] = {x, y};
const float deriv[2][2] = {{dx[0], dx[1]}, {dy[0], dy[1]}};
buffer_->readEWA(output, uv, deriv);
}
} // namespace blender::compositor

28
source/blender/compositor/intern/COM_BufferOperation.h
View File

@ -1,28 +0,0 @@
/* SPDX-FileCopyrightText: 2021 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#include "COM_ConstantOperation.h"
namespace blender::compositor {
class BufferOperation : public ConstantOperation {
private:
MemoryBuffer *buffer_;
MemoryBuffer *inflated_buffer_;
public:
BufferOperation(MemoryBuffer *buffer, DataType data_type);
const float *get_constant_elem() override;
void *initialize_tile_data(rcti *rect) override;
void init_execution() override;
void deinit_execution() override;
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void execute_pixel_filtered(
float output[4], float x, float y, float dx[2], float dy[2]) override;
};
} // namespace blender::compositor

19
source/blender/compositor/intern/COM_CPUDevice.cc
View File

@ -4,8 +4,7 @@
#include "COM_CPUDevice.h"
#include "COM_ExecutionGroup.h"
#include "COM_NodeOperation.h"
#include "COM_WorkPackage.h"
namespace blender::compositor {
@ -13,21 +12,7 @@ CPUDevice::CPUDevice(int thread_id) : thread_id_(thread_id) {}
void CPUDevice::execute(WorkPackage *work_package)
{
switch (work_package->type) {
case eWorkPackageType::Tile: {
const uint chunk_number = work_package->chunk_number;
ExecutionGroup *execution_group = work_package->execution_group;
execution_group->get_output_operation()->execute_region(&work_package->rect, chunk_number);
execution_group->finalize_chunk_execution(chunk_number, nullptr);
break;
}
case eWorkPackageType::CustomFunction: {
work_package->execute_fn();
break;
}
}
work_package->execute_fn();
if (work_package->executed_fn) {
work_package->executed_fn();
}

28
source/blender/compositor/intern/COM_ChunkOrder.cc
View File

@ -1,28 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include <cfloat>
#include "COM_ChunkOrder.h"
namespace blender::compositor {
void ChunkOrder::update_distance(ChunkOrderHotspot *hotspots, uint len_hotspots)
{
double new_distance = DBL_MAX;
for (int index = 0; index < len_hotspots; index++) {
double distance_to_hotspot = hotspots[index].calc_distance(x, y);
if (distance_to_hotspot < new_distance) {
new_distance = distance_to_hotspot;
}
}
this->distance = new_distance;
}
bool operator<(const ChunkOrder &a, const ChunkOrder &b)
{
return a.distance < b.distance;
}
} // namespace blender::compositor

33
source/blender/compositor/intern/COM_ChunkOrder.h
View File

@ -1,33 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#ifdef WITH_CXX_GUARDEDALLOC
# include "MEM_guardedalloc.h"
#endif
#include "BLI_sys_types.h"
#include "COM_ChunkOrderHotspot.h"
namespace blender::compositor {
/** Helper to determine the order how chunks are prioritized during execution. */
struct ChunkOrder {
uint index = 0;
int x = 0;
int y = 0;
double distance = 0.0;
friend bool operator<(const ChunkOrder &a, const ChunkOrder &b);
void update_distance(ChunkOrderHotspot *hotspots, uint len_hotspots);
#ifdef WITH_CXX_GUARDEDALLOC
MEM_CXX_CLASS_ALLOC_FUNCS("COM:ChunkOrderHotspot")
#endif
};
} // namespace blender::compositor

19
source/blender/compositor/intern/COM_ChunkOrderHotspot.cc
View File

@ -1,19 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_ChunkOrderHotspot.h"
#include <cmath>
namespace blender::compositor {
double ChunkOrderHotspot::calc_distance(int x, int y)
{
int dx = this->x - x;
int dy = this->y - y;
double result = sqrt(double(dx * dx + dy * dy));
result += double(this->addition);
return result;
}
} // namespace blender::compositor

27
source/blender/compositor/intern/COM_ChunkOrderHotspot.h
View File

@ -1,27 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#ifdef WITH_CXX_GUARDEDALLOC
# include "MEM_guardedalloc.h"
#endif
namespace blender::compositor {
struct ChunkOrderHotspot {
int x;
int y;
float addition;
ChunkOrderHotspot(int x, int y, float addition) : x(x), y(y), addition(addition) {}
double calc_distance(int x, int y);
#ifdef WITH_CXX_GUARDEDALLOC
MEM_CXX_CLASS_ALLOC_FUNCS("COM:ChunkOrderHotspot")
#endif
};
} // namespace blender::compositor

6
source/blender/compositor/intern/COM_CompositorContext.cc
View File

@ -11,7 +11,6 @@ CompositorContext::CompositorContext()
scene_ = nullptr;
rd_ = nullptr;
quality_ = eCompositorQuality::High;
hasActiveOpenCLDevices_ = false;
fast_calculation_ = false;
bnodetree_ = nullptr;
}
@ -28,9 +27,4 @@ Size2f CompositorContext::get_render_size() const
get_render_data()->ysch * get_render_percentage_as_factor()};
}
eExecutionModel CompositorContext::get_execution_model() const
{
return eExecutionModel::FullFrame;
}
} // namespace blender::compositor

35
source/blender/compositor/intern/COM_CompositorContext.h
View File

@ -59,11 +59,6 @@ class CompositorContext {
*/
bNodeInstanceHash *previews_;
/**
* \brief does this system have active opencl devices?
*/
bool hasActiveOpenCLDevices_;
/**
* \brief Skip slow nodes
*/
@ -180,22 +175,6 @@ class CompositorContext {
*/
int get_framenumber() const;
/**
* \brief has this system active opencl_devices?
*/
bool get_has_active_opencl_devices() const
{
return hasActiveOpenCLDevices_;
}
/**
* \brief set has this system active opencl_devices?
*/
void setHasActiveOpenCLDevices(bool hasAvtiveOpenCLDevices)
{
hasActiveOpenCLDevices_ = hasAvtiveOpenCLDevices;
}
/** Whether it has a view with a specific name and not the default one. */
bool has_explicit_view() const
{
@ -234,11 +213,6 @@ class CompositorContext {
view_name_ = view_name;
}
int get_chunksize() const
{
return 256;
}
void set_fast_calculation(bool fast_calculation)
{
fast_calculation_ = fast_calculation;
@ -247,10 +221,6 @@ class CompositorContext {
{
return fast_calculation_;
}
bool is_groupnode_buffer_enabled() const
{
return false;
}
/**
* \brief Get the render percentage as a factor.
@ -262,11 +232,6 @@ class CompositorContext {
}
Size2f get_render_size() const;
/**
* Get active execution model.
*/
eExecutionModel get_execution_model() const;
};
} // namespace blender::compositor

2
source/blender/compositor/intern/COM_ConstantFolder.cc
View File

@ -133,7 +133,7 @@ Vector<ConstantOperation *> ConstantFolder::try_fold_operations(Span<NodeOperati
int ConstantFolder::fold_operations()
{
WorkScheduler::start(operations_builder_.context());
WorkScheduler::start();
Vector<ConstantOperation *> last_folds = try_fold_operations(
operations_builder_.get_operations());
int folds_count = last_folds.size();

12
source/blender/compositor/intern/COM_Converter.cc
View File

@ -550,13 +550,11 @@ void COM_convert_canvas(NodeOperationBuilder &builder,
builder.add_operation(syop);
rcti scale_canvas = from_operation->get_canvas();
if (builder.context().get_execution_model() == eExecutionModel::FullFrame) {
ScaleOperation::scale_area(scale_canvas, scaleX, scaleY);
scale_canvas.xmax = scale_canvas.xmin + to_operation->get_width();
scale_canvas.ymax = scale_canvas.ymin + to_operation->get_height();
addX = 0;
addY = 0;
}
ScaleOperation::scale_area(scale_canvas, scaleX, scaleY);
scale_canvas.xmax = scale_canvas.xmin + to_operation->get_width();
scale_canvas.ymax = scale_canvas.ymin + to_operation->get_height();
addX = 0;
addY = 0;
scale_operation->set_canvas(scale_canvas);
sxop->set_canvas(scale_canvas);
syop->set_canvas(scale_canvas);

101
source/blender/compositor/intern/COM_Debug.cc
View File

@ -12,11 +12,8 @@
#include "IMB_imbuf.hh"
#include "IMB_imbuf_types.hh"
#include "COM_ExecutionGroup.h"
#include "COM_ReadBufferOperation.h"
#include "COM_SetValueOperation.h"
#include "COM_ViewerOperation.h"
#include "COM_WriteBufferOperation.h"
namespace blender::compositor {
@ -25,7 +22,6 @@ DebugInfo::NodeNameMap DebugInfo::node_names_;
DebugInfo::OpNameMap DebugInfo::op_names_;
std::string DebugInfo::current_node_name_;
std::string DebugInfo::current_op_name_;
DebugInfo::GroupStateMap DebugInfo::group_states_;
static std::string operation_class_name(const NodeOperation *op)
{
@ -56,7 +52,6 @@ std::string DebugInfo::operation_name(const NodeOperation *op)
int DebugInfo::graphviz_operation(const ExecutionSystem *system,
NodeOperation *operation,
const ExecutionGroup *group,
char *str,
int maxlen)
{
@ -78,20 +73,9 @@ int DebugInfo::graphviz_operation(const ExecutionSystem *system,
else if (operation->get_flags().is_set_operation) {
fillcolor = "khaki1";
}
else if (operation->get_flags().is_read_buffer_operation) {
fillcolor = "darkolivegreen3";
}
else if (operation->get_flags().is_write_buffer_operation) {
fillcolor = "darkorange";
}
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "// OPERATION: %p\r\n", operation);
if (group) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "\"O_%p_%p\"", operation, group);
}
else {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "\"O_%p\"", operation);
}
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "\"O_%p\"", operation);
len += snprintf(str + len,
maxlen > len ? maxlen - len : 0,
" [fillcolor=%s,style=filled,shape=record,label=\"{",
@ -220,14 +204,11 @@ int DebugInfo::graphviz_legend_group(
return len;
}
int DebugInfo::graphviz_legend(char *str, int maxlen, const bool has_execution_groups)
int DebugInfo::graphviz_legend(char *str, int maxlen)
{
int len = 0;
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "{\r\n");
if (has_execution_groups) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "rank = sink;\r\n");
}
len += snprintf(
str + len, maxlen > len ? maxlen - len : 0, "Legend [shape=none, margin=0, label=<\r\n");
@ -247,25 +228,9 @@ int DebugInfo::graphviz_legend(char *str, int maxlen, const bool has_execution_g
"Viewer", "lightskyblue3", str + len, maxlen > len ? maxlen - len : 0);
len += graphviz_legend_color(
"Active Viewer", "lightskyblue1", str + len, maxlen > len ? maxlen - len : 0);
if (has_execution_groups) {
len += graphviz_legend_color(
"Write Buffer", "darkorange", str + len, maxlen > len ? maxlen - len : 0);
len += graphviz_legend_color(
"Read Buffer", "darkolivegreen3", str + len, maxlen > len ? maxlen - len : 0);
}
len += graphviz_legend_color(
"Input Value", "khaki1", str + len, maxlen > len ? maxlen - len : 0);
if (has_execution_groups) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "<TR><TD></TD></TR>\r\n");
len += graphviz_legend_group(
"Group Waiting", "white", "dashed", str + len, maxlen > len ? maxlen - len : 0);
len += graphviz_legend_group(
"Group Running", "firebrick1", "solid", str + len, maxlen > len ? maxlen - len : 0);
len += graphviz_legend_group(
"Group Finished", "chartreuse4", "solid", str + len, maxlen > len ? maxlen - len : 0);
}
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "</TABLE>\r\n");
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, ">];\r\n");
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "}\r\n");
@ -275,7 +240,6 @@ int DebugInfo::graphviz_legend(char *str, int maxlen, const bool has_execution_g
bool DebugInfo::graphviz_system(const ExecutionSystem *system, char *str, int maxlen)
{
char strbuf[64];
int len = 0;
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "digraph compositorexecution {\r\n");
@ -284,39 +248,7 @@ bool DebugInfo::graphviz_system(const ExecutionSystem *system, char *str, int ma
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "splines=false\r\n");
std::map<NodeOperation *, std::vector<std::string>> op_groups;
int index = 0;
for (const ExecutionGroup *group : system->groups_) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "// GROUP: %d\r\n", index);
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "subgraph cluster_%d{\r\n", index);
/* used as a check for executing group */
if (group_states_[group] == EG_WAIT) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "style=dashed\r\n");
}
else if (group_states_[group] == EG_RUNNING) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "style=filled\r\n");
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "color=black\r\n");
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "fillcolor=firebrick1\r\n");
}
else if (group_states_[group] == EG_FINISHED) {
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "style=filled\r\n");
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "color=black\r\n");
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "fillcolor=chartreuse4\r\n");
}
for (NodeOperation *operation : group->operations_) {
SNPRINTF(strbuf, "_%p", group);
op_groups[operation].push_back(std::string(strbuf));
len += graphviz_operation(
system, operation, group, str + len, maxlen > len ? maxlen - len : 0);
}
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "}\r\n");
index++;
}
/* operations not included in any group */
for (NodeOperation *operation : system->operations_) {
if (op_groups.find(operation) != op_groups.end()) {
continue;
@ -324,29 +256,7 @@ bool DebugInfo::graphviz_system(const ExecutionSystem *system, char *str, int ma
op_groups[operation].push_back(std::string(""));
len += graphviz_operation(
system, operation, nullptr, str + len, maxlen > len ? maxlen - len : 0);
}
for (NodeOperation *operation : system->operations_) {
if (operation->get_flags().is_read_buffer_operation) {
ReadBufferOperation *read = (ReadBufferOperation *)operation;
WriteBufferOperation *write = read->get_memory_proxy()->get_write_buffer_operation();
std::vector<std::string> &read_groups = op_groups[read];
std::vector<std::string> &write_groups = op_groups[write];
for (int k = 0; k < write_groups.size(); k++) {
for (int l = 0; l < read_groups.size(); l++) {
len += snprintf(str + len,
maxlen > len ? maxlen - len : 0,
"\"O_%p%s\" -> \"O_%p%s\" [style=dotted]\r\n",
write,
write_groups[k].c_str(),
read,
read_groups[l].c_str());
}
}
}
len += graphviz_operation(system, operation, str + len, maxlen > len ? maxlen - len : 0);
}
for (NodeOperation *op : system->operations_) {
@ -401,10 +311,7 @@ bool DebugInfo::graphviz_system(const ExecutionSystem *system, char *str, int ma
}
}
const bool has_execution_groups = system->get_context().get_execution_model() ==
eExecutionModel::Tiled &&
system->groups_.size() > 0;
len += graphviz_legend(str + len, maxlen > len ? maxlen - len : 0, has_execution_groups);
len += graphviz_legend(str + len, maxlen > len ? maxlen - len : 0);
len += snprintf(str + len, maxlen > len ? maxlen - len : 0, "}\r\n");

28
source/blender/compositor/intern/COM_Debug.h
View File

@ -24,15 +24,11 @@ static constexpr bool COM_EXPORT_OPERATION_BUFFERS = false;
class Node;
class NodeOperation;
class ExecutionSystem;
class ExecutionGroup;
class DebugInfo {
public:
typedef enum { EG_WAIT, EG_RUNNING, EG_FINISHED } GroupState;
typedef std::map<const Node *, std::string> NodeNameMap;
typedef std::map<const NodeOperation *, std::string> OpNameMap;
typedef std::map<const ExecutionGroup *, GroupState> GroupStateMap;
static std::string node_name(const Node *node);
static std::string operation_name(const NodeOperation *op);
@ -47,8 +43,6 @@ class DebugInfo {
static std::string current_node_name_;
/** Base name for automatic sub-operations. */
static std::string current_op_name_;
/** For visualizing group states. */
static GroupStateMap group_states_;
public:
static void convert_started()
@ -58,14 +52,10 @@ class DebugInfo {
}
}
static void execute_started(const ExecutionSystem *system)
static void execute_started()
{
if (COM_EXPORT_GRAPHVIZ) {
file_index_ = 1;
group_states_.clear();
for (ExecutionGroup *execution_group : system->groups_) {
group_states_[execution_group] = EG_WAIT;
}
}
if (COM_EXPORT_OPERATION_BUFFERS) {
delete_operation_exports();
@ -100,19 +90,6 @@ class DebugInfo {
}
};
static void execution_group_started(const ExecutionGroup *group)
{
if (COM_EXPORT_GRAPHVIZ) {
group_states_[group] = EG_RUNNING;
}
};
static void execution_group_finished(const ExecutionGroup *group)
{
if (COM_EXPORT_GRAPHVIZ) {
group_states_[group] = EG_FINISHED;
}
};
static void operation_rendered(const NodeOperation *op, MemoryBuffer *render)
{
/* Don't export constant operations as there are too many and it's rarely useful. */
@ -126,7 +103,6 @@ class DebugInfo {
protected:
static int graphviz_operation(const ExecutionSystem *system,
NodeOperation *operation,
const ExecutionGroup *group,
char *str,
int maxlen);
static int graphviz_legend_color(const char *name, const char *color, char *str, int maxlen);
@ -134,7 +110,7 @@ class DebugInfo {
const char *name, const char *color, const char *style, char *str, int maxlen);
static int graphviz_legend_group(
const char *name, const char *color, const char *style, char *str, int maxlen);
static int graphviz_legend(char *str, int maxlen, bool has_execution_groups);
static int graphviz_legend(char *str, int maxlen);
static bool graphviz_system(const ExecutionSystem *system, char *str, int maxlen);
static void export_operation(const NodeOperation *op, MemoryBuffer *render);

19
source/blender/compositor/intern/COM_Enums.cc
View File

@ -43,23 +43,4 @@ std::ostream &operator<<(std::ostream &os, const eCompositorPriority &priority)
return os;
}
std::ostream &operator<<(std::ostream &os, const eWorkPackageState &execution_state)
{
switch (execution_state) {
case eWorkPackageState::NotScheduled: {
os << "ExecutionState::NotScheduled";
break;
}
case eWorkPackageState::Scheduled: {
os << "ExecutionState::Scheduled";
break;
}
case eWorkPackageState::Executed: {
os << "ExecutionState::Executed";
break;
}
}
return os;
}
} // namespace blender::compositor

38
source/blender/compositor/intern/COM_Enums.h
View File

@ -31,48 +31,11 @@ enum class eCompositorQuality {
* \ingroup Execution
*/
enum class eCompositorPriority {
/** \brief High quality setting */
High = 2,
/** \brief Medium quality setting */
Medium = 1,
/** \brief Low quality setting */
Low = 0,
};
/**
* \brief the execution state of a chunk in an ExecutionGroup
* \ingroup Execution
*/
enum class eWorkPackageState {
/**
* \brief chunk is not yet scheduled
*/
NotScheduled = 0,
/**
* \brief chunk is scheduled, but not yet executed
*/
Scheduled = 1,
/**
* \brief chunk is executed.
*/
Executed = 2,
};
/**
* \brief Work type to execute.
* \ingroup Execution
*/
enum class eWorkPackageType {
/**
* \brief Executes an execution group tile.
*/
Tile = 0,
/**
* \brief Executes a custom function.
*/
CustomFunction = 1
};
enum class PixelSampler {
Nearest = 0,
Bilinear = 1,
@ -81,6 +44,5 @@ enum class PixelSampler {
void expand_area_for_sampler(rcti &area, PixelSampler sampler);
std::ostream &operator<<(std::ostream &os, const eCompositorPriority &priority);
std::ostream &operator<<(std::ostream &os, const eWorkPackageState &execution_state);
} // namespace blender::compositor

583
source/blender/compositor/intern/COM_ExecutionGroup.cc
View File

@ -1,583 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_ExecutionGroup.h"
#include "COM_ChunkOrder.h"
#include "COM_Debug.h"
#include "COM_ReadBufferOperation.h"
#include "COM_ViewerOperation.h"
#include "COM_WorkScheduler.h"
#include "COM_WriteBufferOperation.h"
#include "COM_defines.h"
#include "BLI_rand.hh"
#include "BLI_string.h"
#include "BLI_time.h"
#include "BLT_translation.hh"
namespace blender::compositor {
std::ostream &operator<<(std::ostream &os, const ExecutionGroupFlags &flags)
{
if (flags.initialized) {
os << "init,";
}
if (flags.is_output) {
os << "output,";
}
if (flags.complex) {
os << "complex,";
}
if (flags.open_cl) {
os << "open_cl,";
}
if (flags.single_threaded) {
os << "single_threaded,";
}
return os;
}
ExecutionGroup::ExecutionGroup(int id)
{
id_ = id;
bTree_ = nullptr;
height_ = 0;
width_ = 0;
max_read_buffer_offset_ = 0;
x_chunks_len_ = 0;
y_chunks_len_ = 0;
chunks_len_ = 0;
chunks_finished_ = 0;
BLI_rcti_init(&viewer_border_, 0, 0, 0, 0);
execution_start_time_ = 0;
}
std::ostream &operator<<(std::ostream &os, const ExecutionGroup &execution_group)
{
os << "ExecutionGroup(id=" << execution_group.get_id();
os << ",flags={" << execution_group.get_flags() << "}";
os << ",operation=" << *execution_group.get_output_operation() << "";
os << ")";
return os;
}
eCompositorPriority ExecutionGroup::get_render_priority()
{
return this->get_output_operation()->get_render_priority();
}
bool ExecutionGroup::can_contain(NodeOperation &operation)
{
if (!flags_.initialized) {
return true;
}
if (operation.get_flags().is_read_buffer_operation) {
return true;
}
if (operation.get_flags().is_write_buffer_operation) {
return false;
}
if (operation.get_flags().is_set_operation) {
return true;
}
/* complex groups don't allow further ops (except read buffer and values, see above) */
if (flags_.complex) {
return false;
}
/* complex ops can't be added to other groups (except their own, which they initialize, see
* above) */
if (operation.get_flags().complex) {
return false;
}
return true;
}
bool ExecutionGroup::add_operation(NodeOperation *operation)
{
if (!can_contain(*operation)) {
return false;
}
if (!operation->get_flags().is_read_buffer_operation &&
!operation->get_flags().is_write_buffer_operation)
{
flags_.complex = operation->get_flags().complex;
flags_.open_cl = operation->get_flags().open_cl;
flags_.single_threaded = operation->get_flags().single_threaded;
flags_.initialized = true;
}
operations_.append(operation);
return true;
}
NodeOperation *ExecutionGroup::get_output_operation() const
{
/* The first operation of the group is always the output operation. */
return this->operations_[0];
}
void ExecutionGroup::init_work_packages()
{
work_packages_.clear();
if (chunks_len_ != 0) {
work_packages_.resize(chunks_len_);
for (uint index = 0; index < chunks_len_; index++) {
work_packages_[index].type = eWorkPackageType::Tile;
work_packages_[index].state = eWorkPackageState::NotScheduled;
work_packages_[index].execution_group = this;
work_packages_[index].chunk_number = index;
determine_chunk_rect(&work_packages_[index].rect, index);
}
}
}
void ExecutionGroup::init_read_buffer_operations()
{
uint max_offset = 0;
for (NodeOperation *operation : operations_) {
if (operation->get_flags().is_read_buffer_operation) {
ReadBufferOperation *read_operation = static_cast<ReadBufferOperation *>(operation);
read_operations_.append(read_operation);
max_offset = std::max(max_offset, read_operation->get_offset());
}
}
max_offset++;
max_read_buffer_offset_ = max_offset;
}
void ExecutionGroup::init_execution()
{
init_number_of_chunks();
init_work_packages();
init_read_buffer_operations();
}
void ExecutionGroup::deinit_execution()
{
work_packages_.clear();
chunks_len_ = 0;
x_chunks_len_ = 0;
y_chunks_len_ = 0;
read_operations_.clear();
bTree_ = nullptr;
}
void ExecutionGroup::determine_resolution(uint resolution[2])
{
NodeOperation *operation = this->get_output_operation();
resolution[0] = operation->get_width();
resolution[1] = operation->get_height();
this->set_resolution(resolution);
BLI_rcti_init(&viewer_border_, 0, width_, 0, height_);
}
void ExecutionGroup::init_number_of_chunks()
{
if (flags_.single_threaded) {
x_chunks_len_ = 1;
y_chunks_len_ = 1;
chunks_len_ = 1;
}
else {
const float chunk_sizef = chunk_size_;
const int border_width = BLI_rcti_size_x(&viewer_border_);
const int border_height = BLI_rcti_size_y(&viewer_border_);
x_chunks_len_ = ceil(border_width / chunk_sizef);
y_chunks_len_ = ceil(border_height / chunk_sizef);
chunks_len_ = x_chunks_len_ * y_chunks_len_;
}
}
blender::Array<uint> ExecutionGroup::get_execution_order() const
{
blender::Array<uint> chunk_order(chunks_len_);
for (int chunk_index = 0; chunk_index < chunks_len_; chunk_index++) {
chunk_order[chunk_index] = chunk_index;
}
NodeOperation *operation = this->get_output_operation();
float centerX = 0.5f;
float centerY = 0.5f;
ChunkOrdering order_type = ChunkOrdering::Default;
if (operation->get_flags().is_viewer_operation) {
ViewerOperation *viewer = (ViewerOperation *)operation;
centerX = viewer->getCenterX();
centerY = viewer->getCenterY();
order_type = viewer->get_chunk_order();
}
const int border_width = BLI_rcti_size_x(&viewer_border_);
const int border_height = BLI_rcti_size_y(&viewer_border_);
int index;
switch (order_type) {
case ChunkOrdering::Random: {
static blender::RandomNumberGenerator rng;
blender::MutableSpan<uint> span = chunk_order.as_mutable_span();
/* Shuffle twice to make it more random. */
rng.shuffle(span);
rng.shuffle(span);
break;
}
case ChunkOrdering::CenterOut: {
ChunkOrderHotspot hotspot(border_width * centerX, border_height * centerY, 0.0f);
blender::Array<ChunkOrder> chunk_orders(chunks_len_);
for (index = 0; index < chunks_len_; index++) {
const WorkPackage &work_package = work_packages_[index];
chunk_orders[index].index = index;
chunk_orders[index].x = work_package.rect.xmin - viewer_border_.xmin;
chunk_orders[index].y = work_package.rect.ymin - viewer_border_.ymin;
chunk_orders[index].update_distance(&hotspot, 1);
}
std::sort(&chunk_orders[0], &chunk_orders[chunks_len_ - 1]);
for (index = 0; index < chunks_len_; index++) {
chunk_order[index] = chunk_orders[index].index;
}
break;
}
case ChunkOrdering::RuleOfThirds: {
uint tx = border_width / 6;
uint ty = border_height / 6;
uint mx = border_width / 2;
uint my = border_height / 2;
uint bx = mx + 2 * tx;
uint by = my + 2 * ty;
float addition = chunks_len_ / COM_RULE_OF_THIRDS_DIVIDER;
ChunkOrderHotspot hotspots[9] = {
ChunkOrderHotspot(mx, my, addition * 0),
ChunkOrderHotspot(tx, my, addition * 1),
ChunkOrderHotspot(bx, my, addition * 2),
ChunkOrderHotspot(bx, by, addition * 3),
ChunkOrderHotspot(tx, ty, addition * 4),
ChunkOrderHotspot(bx, ty, addition * 5),
ChunkOrderHotspot(tx, by, addition * 6),
ChunkOrderHotspot(mx, ty, addition * 7),
ChunkOrderHotspot(mx, by, addition * 8),
};
blender::Array<ChunkOrder> chunk_orders(chunks_len_);
for (index = 0; index < chunks_len_; index++) {
const WorkPackage &work_package = work_packages_[index];
chunk_orders[index].index = index;
chunk_orders[index].x = work_package.rect.xmin - viewer_border_.xmin;
chunk_orders[index].y = work_package.rect.ymin - viewer_border_.ymin;
chunk_orders[index].update_distance(hotspots, 9);
}
std::sort(&chunk_orders[0], &chunk_orders[chunks_len_]);
for (index = 0; index < chunks_len_; index++) {
chunk_order[index] = chunk_orders[index].index;
}
break;
}
case ChunkOrdering::TopDown:
default:
break;
}
return chunk_order;
}
void ExecutionGroup::execute(ExecutionSystem *graph)
{
const CompositorContext &context = graph->get_context();
const bNodeTree *bTree = context.get_bnodetree();
if (width_ == 0 || height_ == 0) {
return;
} /** \note Break out... no pixels to calculate. */
if (bTree->runtime->test_break && bTree->runtime->test_break(bTree->runtime->tbh)) {
return;
} /** \note Early break out for blur and preview nodes. */
if (chunks_len_ == 0) {
return;
} /** \note Early break out. */
uint chunk_index;
execution_start_time_ = BLI_time_now_seconds();
chunks_finished_ = 0;
bTree_ = bTree;
blender::Array<uint> chunk_order = get_execution_order();
DebugInfo::execution_group_started(this);
DebugInfo::graphviz(graph);
bool breaked = false;
bool finished = false;
uint start_index = 0;
const int max_number_evaluated = BLI_system_thread_count() * 2;
while (!finished && !breaked) {
bool start_evaluated = false;
finished = true;
int number_evaluated = 0;
for (int index = start_index; index < chunks_len_ && number_evaluated < max_number_evaluated;
index++)
{
chunk_index = chunk_order[index];
int y_chunk = chunk_index / x_chunks_len_;
int x_chunk = chunk_index - (y_chunk * x_chunks_len_);
const WorkPackage &work_package = work_packages_[chunk_index];
switch (work_package.state) {
case eWorkPackageState::NotScheduled: {
schedule_chunk_when_possible(graph, x_chunk, y_chunk);
finished = false;
start_evaluated = true;
number_evaluated++;
if (bTree->runtime->update_draw) {
bTree->runtime->update_draw(bTree->runtime->udh);
}
break;
}
case eWorkPackageState::Scheduled: {
finished = false;
start_evaluated = true;
number_evaluated++;
break;
}
case eWorkPackageState::Executed: {
if (!start_evaluated) {
start_index = index + 1;
}
}
};
}
WorkScheduler::finish();
if (bTree->runtime->test_break && bTree->runtime->test_break(bTree->runtime->tbh)) {
breaked = true;
}
}
DebugInfo::execution_group_finished(this);
DebugInfo::graphviz(graph);
}
MemoryBuffer **ExecutionGroup::get_input_buffers_opencl(int chunk_number)
{
WorkPackage &work_package = work_packages_[chunk_number];
MemoryBuffer **memory_buffers = (MemoryBuffer **)MEM_callocN(
sizeof(MemoryBuffer *) * max_read_buffer_offset_, __func__);
rcti output;
for (ReadBufferOperation *read_operation : read_operations_) {
MemoryProxy *memory_proxy = read_operation->get_memory_proxy();
this->determine_depending_area_of_interest(&work_package.rect, read_operation, &output);
MemoryBuffer *memory_buffer =
memory_proxy->get_executor()->construct_consolidated_memory_buffer(*memory_proxy, output);
memory_buffers[read_operation->get_offset()] = memory_buffer;
}
return memory_buffers;
}
MemoryBuffer *ExecutionGroup::construct_consolidated_memory_buffer(MemoryProxy &memory_proxy,
rcti &rect)
{
MemoryBuffer *image_buffer = memory_proxy.get_buffer();
MemoryBuffer *result = new MemoryBuffer(&memory_proxy, rect, MemoryBufferState::Temporary);
result->fill_from(*image_buffer);
return result;
}
void ExecutionGroup::finalize_chunk_execution(int chunk_number, MemoryBuffer **memory_buffers)
{
WorkPackage &work_package = work_packages_[chunk_number];
if (work_package.state == eWorkPackageState::Scheduled) {
work_package.state = eWorkPackageState::Executed;
}
atomic_add_and_fetch_u(&chunks_finished_, 1);
if (memory_buffers) {
for (uint index = 0; index < max_read_buffer_offset_; index++) {
MemoryBuffer *buffer = memory_buffers[index];
if (buffer) {
if (buffer->is_temporarily()) {
memory_buffers[index] = nullptr;
delete buffer;
}
}
}
MEM_freeN(memory_buffers);
}
if (bTree_) {
/* Status report is only performed for top level Execution Groups. */
float progress = chunks_finished_;
progress /= chunks_len_;
bTree_->runtime->progress(bTree_->runtime->prh, progress);
char buf[128];
SNPRINTF(buf, RPT_("Compositing | Tile %u-%u"), chunks_finished_, chunks_len_);
bTree_->runtime->stats_draw(bTree_->runtime->sdh, buf);
}
}
inline void ExecutionGroup::determine_chunk_rect(rcti *r_rect,
const uint x_chunk,
const uint y_chunk) const
{
const int border_width = BLI_rcti_size_x(&viewer_border_);
const int border_height = BLI_rcti_size_y(&viewer_border_);
if (flags_.single_threaded) {
BLI_rcti_init(r_rect, viewer_border_.xmin, border_width, viewer_border_.ymin, border_height);
}
else {
const uint minx = x_chunk * chunk_size_ + viewer_border_.xmin;
const uint miny = y_chunk * chunk_size_ + viewer_border_.ymin;
const uint width = std::min(uint(viewer_border_.xmax), width_);
const uint height = std::min(uint(viewer_border_.ymax), height_);
BLI_rcti_init(r_rect,
std::min(minx, width_),
std::min(minx + chunk_size_, width),
std::min(miny, height_),
std::min(miny + chunk_size_, height));
}
}
void ExecutionGroup::determine_chunk_rect(rcti *r_rect, const uint chunk_number) const
{
const uint y_chunk = chunk_number / x_chunks_len_;
const uint x_chunk = chunk_number - (y_chunk * x_chunks_len_);
determine_chunk_rect(r_rect, x_chunk, y_chunk);
}
MemoryBuffer *ExecutionGroup::allocate_output_buffer(rcti &rect)
{
/* We assume that this method is only called from complex execution groups. */
NodeOperation *operation = this->get_output_operation();
if (operation->get_flags().is_write_buffer_operation) {
WriteBufferOperation *write_operation = (WriteBufferOperation *)operation;
MemoryBuffer *buffer = new MemoryBuffer(
write_operation->get_memory_proxy(), rect, MemoryBufferState::Temporary);
return buffer;
}
return nullptr;
}
bool ExecutionGroup::schedule_area_when_possible(ExecutionSystem *graph, rcti *area)
{
if (flags_.single_threaded) {
return schedule_chunk_when_possible(graph, 0, 0);
}
/* Find all chunks inside the rect
* determine `minxchunk`, `minychunk`, `maxxchunk`, `maxychunk`
* where x and y are chunk-numbers. */
int indexx, indexy;
int minx = max_ii(area->xmin - viewer_border_.xmin, 0);
int maxx = min_ii(area->xmax - viewer_border_.xmin, viewer_border_.xmax - viewer_border_.xmin);
int miny = max_ii(area->ymin - viewer_border_.ymin, 0);
int maxy = min_ii(area->ymax - viewer_border_.ymin, viewer_border_.ymax - viewer_border_.ymin);
int minxchunk = minx / int(chunk_size_);
int maxxchunk = (maxx + int(chunk_size_) - 1) / int(chunk_size_);
int minychunk = miny / int(chunk_size_);
int maxychunk = (maxy + int(chunk_size_) - 1) / int(chunk_size_);
minxchunk = max_ii(minxchunk, 0);
minychunk = max_ii(minychunk, 0);
maxxchunk = min_ii(maxxchunk, int(x_chunks_len_));
maxychunk = min_ii(maxychunk, int(y_chunks_len_));
bool result = true;
for (indexx = minxchunk; indexx < maxxchunk; indexx++) {
for (indexy = minychunk; indexy < maxychunk; indexy++) {
if (!schedule_chunk_when_possible(graph, indexx, indexy)) {
result = false;
}
}
}
return result;
}
bool ExecutionGroup::schedule_chunk(uint chunk_number)
{
WorkPackage &work_package = work_packages_[chunk_number];
if (work_package.state == eWorkPackageState::NotScheduled) {
work_package.state = eWorkPackageState::Scheduled;
WorkScheduler::schedule(&work_package);
return true;
}
return false;
}
bool ExecutionGroup::schedule_chunk_when_possible(ExecutionSystem *graph,
const int chunk_x,
const int chunk_y)
{
if (chunk_x < 0 || chunk_x >= int(x_chunks_len_)) {
return true;
}
if (chunk_y < 0 || chunk_y >= int(y_chunks_len_)) {
return true;
}
/* Check if chunk is already executed or scheduled and not yet executed. */
const int chunk_index = chunk_y * x_chunks_len_ + chunk_x;
WorkPackage &work_package = work_packages_[chunk_index];
if (work_package.state == eWorkPackageState::Executed) {
return true;
}
if (work_package.state == eWorkPackageState::Scheduled) {
return false;
}
bool can_be_executed = true;
rcti area;
for (ReadBufferOperation *read_operation : read_operations_) {
BLI_rcti_init(&area, 0, 0, 0, 0);
MemoryProxy *memory_proxy = read_operation->get_memory_proxy();
determine_depending_area_of_interest(&work_package.rect, read_operation, &area);
ExecutionGroup *group = memory_proxy->get_executor();
if (!group->schedule_area_when_possible(graph, &area)) {
can_be_executed = false;
}
}
if (can_be_executed) {
schedule_chunk(chunk_index);
}
return false;
}
void ExecutionGroup::determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output)
{
this->get_output_operation()->determine_depending_area_of_interest(
input, read_operation, output);
}
void ExecutionGroup::set_viewer_border(float xmin, float xmax, float ymin, float ymax)
{
const NodeOperation &operation = *this->get_output_operation();
if (operation.get_flags().use_viewer_border) {
BLI_rcti_init(&viewer_border_, xmin * width_, xmax * width_, ymin * height_, ymax * height_);
}
}
void ExecutionGroup::set_render_border(float xmin, float xmax, float ymin, float ymax)
{
const NodeOperation &operation = *this->get_output_operation();
if (operation.is_output_operation(true) && operation.get_flags().use_render_border) {
BLI_rcti_init(&viewer_border_, xmin * width_, xmax * width_, ymin * height_, ymax * height_);
}
}
} // namespace blender::compositor

392
source/blender/compositor/intern/COM_ExecutionGroup.h
View File

@ -1,392 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#ifdef WITH_CXX_GUARDEDALLOC
# include "MEM_guardedalloc.h"
#endif
#include <iostream>
#include "BLI_array.hh"
#include "BLI_vector.hh"
#include "COM_Enums.h"
#include "COM_WorkPackage.h"
#include "DNA_node_types.h"
#include "DNA_vec_types.h"
namespace blender::compositor {
class ExecutionSystem;
class NodeOperation;
class MemoryProxy;
class MemoryBuffer;
class ReadBufferOperation;
struct ExecutionGroupFlags {
bool initialized : 1;
/**
* Is this ExecutionGroup an output ExecutionGroup
* An OutputExecution group are groups containing a
* ViewerOperation, CompositeOperation, PreviewOperation.
*/
bool is_output : 1;
bool complex : 1;
/**
* Can this ExecutionGroup be scheduled on an OpenCLDevice.
*/
bool open_cl : 1;
/**
* Schedule this execution group as a single chunk. This
* chunk will be executed by a single thread.
*/
bool single_threaded : 1;
ExecutionGroupFlags()
{
initialized = false;
is_output = false;
complex = false;
open_cl = false;
single_threaded = false;
}
};
std::ostream &operator<<(std::ostream &os, const ExecutionGroupFlags &flags);
/**
* \brief Class ExecutionGroup is a group of Operations that are executed as one.
* This grouping is used to combine Operations that can be executed as one whole when
* multi-processing.
* \ingroup Execution
*/
class ExecutionGroup {
private:
// fields
/**
* Id of the execution group. For debugging purposes.
*/
int id_;
/**
* \brief list of operations in this ExecutionGroup
*/
Vector<NodeOperation *> operations_;
ExecutionGroupFlags flags_;
/**
* \brief Width of the output
*/
unsigned int width_;
/**
* \brief Height of the output
*/
unsigned int height_;
/**
* \brief size of a single chunk, being Width or of height
* a chunk is always a square, except at the edges of the MemoryBuffer
*/
unsigned int chunk_size_;
/**
* \brief number of chunks in the x-axis
*/
unsigned int x_chunks_len_;
/**
* \brief number of chunks in the y-axis
*/
unsigned int y_chunks_len_;
/**
* \brief total number of chunks
*/
unsigned int chunks_len_;
/**
* \brief what is the maximum number field of all ReadBufferOperation in this ExecutionGroup.
* \note this is used to construct the MemoryBuffers that will be passed during execution.
*/
unsigned int max_read_buffer_offset_;
/**
* \brief All read operations of this execution group.
*/
Vector<ReadBufferOperation *> read_operations_;
/**
* \brief reference to the original bNodeTree,
* this field is only set for the 'top' execution group.
* \note can only be used to call the callbacks for progress, status and break.
*/
const bNodeTree *bTree_;
/**
* \brief total number of chunks that have been calculated for this ExecutionGroup
*/
unsigned int chunks_finished_;
/**
* \brief work_packages_ holds all unit of work.
*/
Vector<WorkPackage> work_packages_;
/**
* \brief denotes boundary for border compositing
* \note measured in pixel space
*/
rcti viewer_border_;
/**
* \brief start time of execution
*/
double execution_start_time_;
// methods
/**
* \brief check whether parameter operation can be added to the execution group
* \param operation: the operation to be added
*/
bool can_contain(NodeOperation &operation);
/**
* \brief Determine the rect (minx, maxx, miny, maxy) of a chunk at a position.
*/
void determine_chunk_rect(rcti *r_rect, unsigned int x_chunk, unsigned int y_chunk) const;
/**
* \brief determine the number of chunks, based on the chunk_size, width and height.
* \note The result are stored in the fields number_of_chunks, number_of_xchunks,
* number_of_ychunks
*/
void init_number_of_chunks();
/**
* \brief try to schedule a specific chunk.
* \note scheduling succeeds when all input requirements are met and the chunks hasn't been
* scheduled yet.
* \param graph:
* \param x_chunk:
* \param y_chunk:
* \return [true:false]
* true: package(s) are scheduled
* false: scheduling is deferred (depending workpackages are scheduled)
*/
bool schedule_chunk_when_possible(ExecutionSystem *graph, int chunk_x, int chunk_y);
/**
* \brief try to schedule a specific area.
* \note Check if a certain area is available, when not available this are will be checked.
* \note This method is called from other ExecutionGroup's.
* \param graph:
* \param area:
* \return [true:false]
* true: package(s) are scheduled
* false: scheduling is deferred (depending workpackages are scheduled)
*/
bool schedule_area_when_possible(ExecutionSystem *graph, rcti *area);
/**
* \brief add a chunk to the WorkScheduler.
* \param chunknumber:
*/
bool schedule_chunk(unsigned int chunk_number);
/**
* \brief determine the area of interest of a certain input area
* \note This method only evaluates a single ReadBufferOperation
* \param input: the input area
* \param read_operation: The ReadBufferOperation where the area needs to be evaluated
* \param output: the area needed of the ReadBufferOperation. Result
*/
void determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output);
/**
* Return the execution order of the user visible chunks.
*/
blender::Array<unsigned int> get_execution_order() const;
void init_read_buffer_operations();
void init_work_packages();
public:
// constructors
ExecutionGroup(int id);
int get_id() const
{
return id_;
}
const ExecutionGroupFlags get_flags() const
{
return flags_;
}
// methods
/**
* \brief add an operation to this ExecutionGroup
* \note this method will add input of the operations recursively
* \note this method can create multiple ExecutionGroup's
* \param system:
* \param operation:
* \return True if the operation was successfully added
*/
bool add_operation(NodeOperation *operation);
/**
* \brief set whether this ExecutionGroup is an output
* \param is_output:
*/
void set_output_execution_group(bool is_output)
{
flags_.is_output = is_output;
}
/**
* \brief determine the resolution of this ExecutionGroup
* \param resolution:
*/
void determine_resolution(unsigned int resolution[2]);
/**
* \brief set the resolution of this executiongroup
* \param resolution:
*/
void set_resolution(unsigned int resolution[2])
{
width_ = resolution[0];
height_ = resolution[1];
}
/**
* \brief get the width of this execution group
*/
unsigned int get_width() const
{
return width_;
}
/**
* \brief get the height of this execution group
*/
unsigned int get_height() const
{
return height_;
}
/**
* \brief get the output operation of this ExecutionGroup
* \return NodeOperation *output operation
*/
NodeOperation *get_output_operation() const;
/**
* \brief compose multiple chunks into a single chunk
* \return `(Memorybuffer *)` consolidated chunk
*/
MemoryBuffer *construct_consolidated_memory_buffer(MemoryProxy &memory_proxy, rcti &rect);
/**
* \brief init_execution is called just before the execution of the whole graph will be done.
* \note The implementation will calculate the chunk_size of this execution group.
*/
void init_execution();
/**
* \brief get all input-buffers needed to calculate an chunk
* \note all input-buffers must be executed
* \param chunk_number: the chunk to be calculated
* \return `(MemoryBuffer **)` the input-buffers.
*/
MemoryBuffer **get_input_buffers_opencl(int chunk_number);
/**
* \brief allocate the output-buffer of a chunk
* \param chunk_number: the number of the chunk in the ExecutionGroup
* \param rect: the rect of that chunk
* \see determine_chunk_rect
*/
MemoryBuffer *allocate_output_buffer(rcti &rect);
/**
* \brief after a chunk is executed the needed resources can be freed or unlocked.
* \param chunknumber:
* \param memorybuffers:
*/
void finalize_chunk_execution(int chunk_number, MemoryBuffer **memory_buffers);
/**
* \brief deinit_execution is called just after execution the whole graph.
* \note It will release all needed resources
*/
void deinit_execution();
/**
* \brief schedule an ExecutionGroup
* \note this method will return when all chunks have been calculated, or the execution has
* breaked (by user)
*
* first the order of the chunks will be determined. This is determined by finding the
* ViewerOperation and get the relevant information from it.
* - ChunkOrdering
* - CenterX
* - CenterY
*
* After determining the order of the chunks the chunks will be scheduled
*
* \see ViewerOperation
* \param graph:
*/
/**
* This method is called for the top execution groups. containing the compositor node or the
* preview node or the viewer node).
*/
void execute(ExecutionSystem *graph);
/**
* \brief Determine the rect (minx, maxx, miny, maxy) of a chunk.
*/
void determine_chunk_rect(rcti *r_rect, unsigned int chunk_number) const;
void set_chunksize(int chunksize)
{
chunk_size_ = chunksize;
}
/**
* \brief get the Render priority of this ExecutionGroup
* \see ExecutionSystem.execute
*/
eCompositorPriority get_render_priority();
/**
* \brief set border for viewer operation
* \note all the coordinates are assumed to be in normalized space
*/
void set_viewer_border(float xmin, float xmax, float ymin, float ymax);
void set_render_border(float xmin, float xmax, float ymin, float ymax);
/* allow the DebugInfo class to look at internals */
friend class DebugInfo;
#ifdef WITH_CXX_GUARDEDALLOC
MEM_CXX_CLASS_ALLOC_FUNCS("COM:ExecutionGroup")
#endif
};
std::ostream &operator<<(std::ostream &os, const ExecutionGroup &execution_group);
} // namespace blender::compositor

27
source/blender/compositor/intern/COM_ExecutionSystem.cc
View File

@ -5,11 +5,9 @@
#include "COM_ExecutionSystem.h"
#include "COM_Debug.h"
#include "COM_ExecutionGroup.h"
#include "COM_FullFrameExecutionModel.h"
#include "COM_NodeOperation.h"
#include "COM_NodeOperationBuilder.h"
#include "COM_TiledExecutionModel.h"
#include "COM_WorkPackage.h"
#include "COM_WorkScheduler.h"
@ -44,7 +42,6 @@ ExecutionSystem::ExecutionSystem(RenderData *rd,
context_.set_quality((eCompositorQuality)editingtree->edit_quality);
}
context_.set_rendering(rendering);
context_.setHasActiveOpenCLDevices(WorkScheduler::has_gpu_devices() && false);
context_.set_render_data(rd);
@ -56,17 +53,7 @@ ExecutionSystem::ExecutionSystem(RenderData *rd,
builder.convert_to_operations(this);
}
switch (context_.get_execution_model()) {
case eExecutionModel::Tiled:
execution_model_ = new TiledExecutionModel(context_, operations_, groups_);
break;
case eExecutionModel::FullFrame:
execution_model_ = new FullFrameExecutionModel(context_, active_buffers_, operations_);
break;
default:
BLI_assert_msg(0, "Non implemented execution model");
break;
}
execution_model_ = new FullFrameExecutionModel(context_, active_buffers_, operations_);
}
ExecutionSystem::~ExecutionSystem()
@ -80,23 +67,16 @@ ExecutionSystem::~ExecutionSystem()
delete operation;
}
operations_.clear();
for (ExecutionGroup *group : groups_) {
delete group;
}
groups_.clear();
}
void ExecutionSystem::set_operations(const Span<NodeOperation *> operations,
const Span<ExecutionGroup *> groups)
void ExecutionSystem::set_operations(const Span<NodeOperation *> operations)
{
operations_ = operations;
groups_ = groups;
}
void ExecutionSystem::execute()
{
DebugInfo::execute_started(this);
DebugInfo::execute_started();
for (NodeOperation *op : operations_) {
op->init_data();
}
@ -131,7 +111,6 @@ void ExecutionSystem::execute_work(const rcti &work_rect,
}
WorkPackage &sub_work = sub_works[i];
sub_work.type = eWorkPackageType::CustomFunction;
sub_work.execute_fn = [=, &work_func, &work_rect]() {
if (is_breaked()) {
return;

45
source/blender/compositor/intern/COM_ExecutionSystem.h
View File

@ -82,36 +82,9 @@ class ProfilerData;
*
* \see COM_convert_data_type Datatype conversions
* \see Converter.convert_resolution Image size conversions
*
* \section EM_Step4 Step4: group operations in executions groups
* ExecutionGroup are groups of operations that are calculated as being one bigger operation.
* All operations will be part of an ExecutionGroup.
* Complex nodes will be added to separate groups. Between ExecutionGroup's the data will be stored
* in MemoryBuffers. ReadBufferOperations and WriteBufferOperations are added where needed.
*
* <pre>
*
* +------------------------------+ +----------------+
* | ExecutionGroup A | |ExecutionGroup B| ExecutionGroup
* | +----------+ +----------+| |+----------+ |
* /----->| Operation|---->| Operation|-\ /--->| Operation|-\ | NodeOperation
* | | | A | | B ||| | || C | | |
* | | | cFFA | /->| cFFA ||| | || cFFA | | |
* | | +----------+ | +----------+|| | |+----------+ | |
* | +---------------|--------------+v | +-------------v--+
* +-*----+ +---*--+ +--*-*--+ +--*----+
* |inputA| |inputB| |outputA| |outputB| MemoryBuffer
* |cFAA | |cFAA | |cFAA | |cFAA |
* +------+ +------+ +-------+ +-------+
* </pre>
* \see ExecutionSystem.group_operations method doing this step
* \see ExecutionSystem.add_read_write_buffer_operations
* \see NodeOperation.is_complex
* \see ExecutionGroup class representing the ExecutionGroup
*/
/* Forward declarations. */
class ExecutionGroup;
class ExecutionModel;
class NodeOperation;
@ -136,11 +109,6 @@ class ExecutionSystem {
*/
Vector<NodeOperation *> operations_;
/**
* \brief vector of groups
*/
Vector<ExecutionGroup *> groups_;
/**
* Active execution model implementation.
*/
@ -178,13 +146,13 @@ class ExecutionSystem {
*/
~ExecutionSystem();
void set_operations(Span<NodeOperation *> operations, Span<ExecutionGroup *> groups);
void set_operations(Span<NodeOperation *> operations);
/**
* \brief execute this system
* - initialize the NodeOperation's and ExecutionGroup's
* - schedule the output ExecutionGroup's based on their priority
* - deinitialize the ExecutionGroup's and NodeOperation's
* - initialize the NodeOperation's
* - schedule the outputs based on their priority
* - deinitialize the NodeOperation's
*/
void execute();
@ -196,11 +164,6 @@ class ExecutionSystem {
return context_;
}
SharedOperationBuffers &get_active_buffers()
{
return active_buffers_;
}
/**
* Multi-threadedly execute given work function passing work_rect splits as argument.
*/

2
source/blender/compositor/intern/COM_FullFrameExecutionModel.cc
View File

@ -134,7 +134,7 @@ void FullFrameExecutionModel::render_operations()
{
const bool is_rendering = context_.is_rendering();
WorkScheduler::start(this->context_);
WorkScheduler::start();
for (eCompositorPriority priority : priorities_) {
for (NodeOperation *op : operations_) {
const bool has_size = op->get_width() > 0 && op->get_height() > 0;

60
source/blender/compositor/intern/COM_MemoryBuffer.cc
View File

@ -4,8 +4,6 @@
#include "COM_MemoryBuffer.h"
#include "COM_MemoryProxy.h"
#include "IMB_colormanagement.hh"
#include "IMB_imbuf_types.hh"
@ -30,31 +28,14 @@ static rcti create_rect(const int width, const int height)
return rect;
}
MemoryBuffer::MemoryBuffer(MemoryProxy *memory_proxy, const rcti &rect, MemoryBufferState state)
{
rect_ = rect;
is_a_single_elem_ = false;
memory_proxy_ = memory_proxy;
num_channels_ = COM_data_type_num_channels(memory_proxy->get_data_type());
buffer_ = (float *)MEM_mallocN_aligned(
sizeof(float) * buffer_len() * num_channels_, 16, "COM_MemoryBuffer");
owns_data_ = true;
state_ = state;
datatype_ = memory_proxy->get_data_type();
set_strides();
}
MemoryBuffer::MemoryBuffer(DataType data_type, const rcti &rect, bool is_a_single_elem)
{
rect_ = rect;
is_a_single_elem_ = is_a_single_elem;
memory_proxy_ = nullptr;
num_channels_ = COM_data_type_num_channels(data_type);
buffer_ = (float *)MEM_mallocN_aligned(
sizeof(float) * buffer_len() * num_channels_, 16, "COM_MemoryBuffer");
owns_data_ = true;
state_ = MemoryBufferState::Temporary;
datatype_ = data_type;
set_strides();
@ -73,19 +54,16 @@ MemoryBuffer::MemoryBuffer(float *buffer,
{
rect_ = rect;
is_a_single_elem_ = is_a_single_elem;
memory_proxy_ = nullptr;
num_channels_ = num_channels;
datatype_ = COM_num_channels_data_type(num_channels);
buffer_ = buffer;
owns_data_ = false;
state_ = MemoryBufferState::Temporary;
set_strides();
}
MemoryBuffer::MemoryBuffer(const MemoryBuffer &src) : MemoryBuffer(src.datatype_, src.rect_, false)
{
memory_proxy_ = src.memory_proxy_;
/* src may be single elem buffer */
fill_from(src);
}
@ -468,44 +446,6 @@ void MemoryBuffer::read_elem_filtered(
out);
}
/* TODO(manzanilla): to be removed with tiled implementation. */
static void read_ewa_pixel_sampled(void *userdata, int x, int y, float result[4])
{
MemoryBuffer *buffer = (MemoryBuffer *)userdata;
buffer->read(result, x, y);
}
/* TODO(manzanilla): to be removed with tiled implementation. */
void MemoryBuffer::readEWA(float *result, const float uv[2], const float derivatives[2][2])
{
if (is_a_single_elem_) {
memcpy(result, buffer_, sizeof(float) * num_channels_);
}
else {
BLI_assert(datatype_ == DataType::Color);
float inv_width = 1.0f / float(this->get_width()),
inv_height = 1.0f / float(this->get_height());
/* TODO(sergey): Render pipeline uses normalized coordinates and derivatives,
* but compositor uses pixel space. For now let's just divide the values and
* switch compositor to normalized space for EWA later.
*/
float uv_normal[2] = {uv[0] * inv_width, uv[1] * inv_height};
float du_normal[2] = {derivatives[0][0] * inv_width, derivatives[0][1] * inv_height};
float dv_normal[2] = {derivatives[1][0] * inv_width, derivatives[1][1] * inv_height};
BLI_ewa_filter(this->get_width(),
this->get_height(),
false,
true,
uv_normal,
du_normal,
dv_normal,
read_ewa_pixel_sampled,
this,
result);
}
}
void MemoryBuffer::copy_single_elem_from(const MemoryBuffer *src,
const int channel_offset,
const int elem_size,

73
source/blender/compositor/intern/COM_MemoryBuffer.h
View File

@ -21,28 +21,14 @@ struct ImBuf;
namespace blender::compositor {
/**
* \brief state of a memory buffer
* \ingroup Memory
*/
enum class MemoryBufferState {
/** \brief memory has been allocated on creator device and CPU machine,
* but kernel has not been executed */
Default = 0,
/** \brief chunk is consolidated from other chunks. special state. */
Temporary = 6,
};
enum class MemoryBufferExtend {
Clip,
Extend,
Repeat,
};
class MemoryProxy;
/**
* \brief a MemoryBuffer contains access to the data of a chunk
* \brief a MemoryBuffer contains access to the data
*/
class MemoryBuffer {
public:
@ -65,26 +51,16 @@ class MemoryBuffer {
int row_stride;
private:
/**
* \brief proxy of the memory (same for all chunks in the same buffer)
*/
MemoryProxy *memory_proxy_;
/**
* \brief the type of buffer DataType::Value, DataType::Vector, DataType::Color
*/
DataType datatype_;
/**
* \brief region of this buffer inside relative to the MemoryProxy
* \brief region of this buffer inside
*/
rcti rect_;
/**
* \brief state of the buffer
*/
MemoryBufferState state_;
/**
* \brief the actual float buffer/data
*/
@ -113,11 +89,6 @@ class MemoryBuffer {
int to_positive_y_stride_;
public:
/**
* \brief construct new temporarily MemoryBuffer for an area
*/
MemoryBuffer(MemoryProxy *memory_proxy, const rcti &rect, MemoryBufferState state);
/**
* \brief construct new temporarily MemoryBuffer for an area
*/
@ -384,12 +355,6 @@ class MemoryBuffer {
return buffer_;
}
float *release_ownership_buffer()
{
owns_data_ = false;
return buffer_;
}
/**
* Converts a single elem buffer to a full size buffer (allocates memory for all
* elements in resolution).
@ -491,8 +456,6 @@ class MemoryBuffer {
y = y + rect_.ymin;
}
/* TODO(manzanilla): to be removed with tiled implementation. For applying #MemoryBufferExtend
* use #wrap_pixel. */
inline void read(float *result,
int x,
int y,
@ -514,28 +477,6 @@ class MemoryBuffer {
memcpy(result, buffer, sizeof(float) * num_channels_);
}
}
/* TODO(manzanilla): to be removed with tiled implementation. */
inline void read_no_check(float *result,
int x,
int y,
MemoryBufferExtend extend_x = MemoryBufferExtend::Clip,
MemoryBufferExtend extend_y = MemoryBufferExtend::Clip)
{
int u = x;
int v = y;
this->wrap_pixel(u, v, extend_x, extend_y);
const int offset = get_coords_offset(u, v);
BLI_assert(offset >= 0);
BLI_assert(offset < this->buffer_len() * num_channels_);
BLI_assert(!(extend_x == MemoryBufferExtend::Clip && (u < rect_.xmin || u >= rect_.xmax)) &&
!(extend_y == MemoryBufferExtend::Clip && (v < rect_.ymin || v >= rect_.ymax)));
float *buffer = &buffer_[offset];
memcpy(result, buffer, sizeof(float) * num_channels_);
}
void write_pixel(int x, int y, const float color[4]);
void add_pixel(int x, int y, const float color[4]);
inline void read_bilinear(float *result,
@ -569,16 +510,6 @@ class MemoryBuffer {
}
}
void readEWA(float *result, const float uv[2], const float derivatives[2][2]);
/**
* \brief is this MemoryBuffer a temporarily buffer (based on an area, not on a chunk)
*/
inline bool is_temporarily() const
{
return state_ == MemoryBufferState::Temporary;
}
/**
* \brief Apply a color processor on the given area.
*/

37
source/blender/compositor/intern/COM_MemoryProxy.cc
View File

@ -1,37 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_MemoryProxy.h"
#include "COM_MemoryBuffer.h"
namespace blender::compositor {
MemoryProxy::MemoryProxy(DataType datatype)
{
write_buffer_operation_ = nullptr;
executor_ = nullptr;
buffer_ = nullptr;
datatype_ = datatype;
}
void MemoryProxy::allocate(uint width, uint height)
{
rcti result;
result.xmin = 0;
result.xmax = width;
result.ymin = 0;
result.ymax = height;
buffer_ = new MemoryBuffer(this, result, MemoryBufferState::Default);
}
void MemoryProxy::free()
{
if (buffer_) {
delete buffer_;
buffer_ = nullptr;
}
}
} // namespace blender::compositor

114
source/blender/compositor/intern/COM_MemoryProxy.h
View File

@ -1,114 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#ifdef WITH_CXX_GUARDEDALLOC
# include "MEM_guardedalloc.h"
#endif
#include "COM_defines.h"
namespace blender::compositor {
/* Forward declarations. */
class MemoryBuffer;
class ExecutionGroup;
class WriteBufferOperation;
/**
* \brief A MemoryProxy is a unique identifier for a memory buffer.
* A single MemoryProxy is used among all chunks of the same buffer,
* the MemoryBuffer only stores the data of a single chunk.
* \ingroup Memory
*/
class MemoryProxy {
private:
/**
* \brief reference to the output operation of the executiongroup
*/
WriteBufferOperation *write_buffer_operation_;
/**
* \brief reference to the executor. the Execution group that can fill a chunk
*/
ExecutionGroup *executor_;
/**
* \brief the allocated memory
*/
MemoryBuffer *buffer_;
/**
* \brief datatype of this MemoryProxy
*/
DataType datatype_;
public:
MemoryProxy(DataType type);
/**
* \brief set the ExecutionGroup that can be scheduled to calculate a certain chunk.
* \param group: the ExecutionGroup to set
*/
void set_executor(ExecutionGroup *executor)
{
executor_ = executor;
}
/**
* \brief get the ExecutionGroup that can be scheduled to calculate a certain chunk.
*/
ExecutionGroup *get_executor() const
{
return executor_;
}
/**
* \brief set the WriteBufferOperation that is responsible for writing to this MemoryProxy
* \param operation:
*/
void set_write_buffer_operation(WriteBufferOperation *operation)
{
write_buffer_operation_ = operation;
}
/**
* \brief get the WriteBufferOperation that is responsible for writing to this MemoryProxy
* \return WriteBufferOperation
*/
WriteBufferOperation *get_write_buffer_operation() const
{
return write_buffer_operation_;
}
/**
* \brief allocate memory of size width x height
*/
void allocate(unsigned int width, unsigned int height);
/**
* \brief free the allocated memory
*/
void free();
/**
* \brief get the allocated memory
*/
inline MemoryBuffer *get_buffer()
{
return buffer_;
}
inline DataType get_data_type()
{
return datatype_;
}
#ifdef WITH_CXX_GUARDEDALLOC
MEM_CXX_CLASS_ALLOC_FUNCS("COM:MemoryProxy")
#endif
};
} // namespace blender::compositor

1
source/blender/compositor/intern/COM_MultiThreadedOperation.cc
View File

@ -11,7 +11,6 @@ MultiThreadedOperation::MultiThreadedOperation()
{
num_passes_ = 1;
current_pass_ = 0;
flags_.is_fullframe_operation = true;
}
void MultiThreadedOperation::update_memory_buffer(MemoryBuffer *output,

14
source/blender/compositor/intern/COM_NodeGraph.cc
View File

@ -231,7 +231,7 @@ void NodeGraph::add_proxies_group_inputs(bNode *b_node, bNode *b_node_io)
}
}
void NodeGraph::add_proxies_group_outputs(const CompositorContext &context,
void NodeGraph::add_proxies_group_outputs(const CompositorContext & /*context*/,
bNode *b_node,
bNode *b_node_io)
{
@ -247,16 +247,8 @@ void NodeGraph::add_proxies_group_outputs(const CompositorContext &context,
{
bNodeSocket *b_sock_group = find_b_node_output(b_node, b_sock_io->identifier);
if (b_sock_group) {
if (context.is_groupnode_buffer_enabled() &&
context.get_execution_model() == eExecutionModel::Tiled)
{
SocketBufferNode *buffer = new SocketBufferNode(b_node_io, b_sock_io, b_sock_group);
add_node(buffer, b_group_tree, key, is_active_group);
}
else {
SocketProxyNode *proxy = new SocketProxyNode(b_node_io, b_sock_io, b_sock_group, true);
add_node(proxy, b_group_tree, key, is_active_group);
}
SocketProxyNode *proxy = new SocketProxyNode(b_node_io, b_sock_io, b_sock_group, true);
add_node(proxy, b_group_tree, key, is_active_group);
}
}
}

189
source/blender/compositor/intern/COM_NodeOperation.cc
View File

@ -4,9 +4,7 @@
#include <cstdio>
#include "COM_BufferOperation.h"
#include "COM_ExecutionSystem.h"
#include "COM_ReadBufferOperation.h"
#include "COM_NodeOperation.h" /* own include */
@ -156,26 +154,6 @@ void NodeOperation::init_execution()
/* pass */
}
void NodeOperation::init_mutex()
{
BLI_mutex_init(&mutex_);
}
void NodeOperation::lock_mutex()
{
BLI_mutex_lock(&mutex_);
}
void NodeOperation::unlock_mutex()
{
BLI_mutex_unlock(&mutex_);
}
void NodeOperation::deinit_mutex()
{
BLI_mutex_end(&mutex_);
}
void NodeOperation::deinit_execution()
{
/* pass */
@ -213,57 +191,15 @@ NodeOperation *NodeOperation::get_input_operation(int index)
return nullptr;
}
bool NodeOperation::determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output)
{
if (inputs_.is_empty()) {
BLI_rcti_init(output, input->xmin, input->xmax, input->ymin, input->ymax);
return false;
}
rcti temp_output;
bool first = true;
for (int i = 0; i < get_number_of_input_sockets(); i++) {
NodeOperation *input_operation = this->get_input_operation(i);
if (input_operation &&
input_operation->determine_depending_area_of_interest(input, read_operation, &temp_output))
{
if (first) {
output->xmin = temp_output.xmin;
output->ymin = temp_output.ymin;
output->xmax = temp_output.xmax;
output->ymax = temp_output.ymax;
first = false;
}
else {
output->xmin = std::min(output->xmin, temp_output.xmin);
output->ymin = std::min(output->ymin, temp_output.ymin);
output->xmax = std::max(output->xmax, temp_output.xmax);
output->ymax = std::max(output->ymax, temp_output.ymax);
}
}
}
return !first;
}
/* -------------------------------------------------------------------- */
/** \name Full Frame Methods
* \{ */
void NodeOperation::get_area_of_interest(const int input_idx,
void NodeOperation::get_area_of_interest(const int /*input_idx*/,
const rcti &output_area,
rcti &r_input_area)
{
if (get_flags().is_fullframe_operation) {
r_input_area = output_area;
}
else {
/* Non full-frame operations never implement this method. To ensure correctness assume
* whole area is used. */
NodeOperation *input_op = get_input_operation(input_idx);
r_input_area = input_op->get_canvas();
}
r_input_area = output_area;
}
void NodeOperation::get_area_of_interest(NodeOperation *input_op,
@ -283,12 +219,7 @@ void NodeOperation::render(MemoryBuffer *output_buf,
Span<rcti> areas,
Span<MemoryBuffer *> inputs_bufs)
{
if (get_flags().is_fullframe_operation) {
render_full_frame(output_buf, areas, inputs_bufs);
}
else {
render_full_frame_fallback(output_buf, areas, inputs_bufs);
}
render_full_frame(output_buf, areas, inputs_bufs);
}
void NodeOperation::render_full_frame(MemoryBuffer *output_buf,
@ -302,92 +233,6 @@ void NodeOperation::render_full_frame(MemoryBuffer *output_buf,
deinit_execution();
}
void NodeOperation::render_full_frame_fallback(MemoryBuffer *output_buf,
Span<rcti> areas,
Span<MemoryBuffer *> inputs_bufs)
{
Vector<NodeOperationOutput *> orig_input_links = replace_inputs_with_buffers(inputs_bufs);
init_execution();
const bool is_output_operation = get_number_of_output_sockets() == 0;
if (!is_output_operation && output_buf->is_a_single_elem()) {
float *output_elem = output_buf->get_elem(0, 0);
read_sampled(output_elem, 0, 0, PixelSampler::Nearest);
}
else {
for (const rcti &rect : areas) {
exec_system_->execute_work(rect, [=](const rcti &split_rect) {
rcti tile_rect = split_rect;
if (is_output_operation) {
execute_region(&tile_rect, 0);
}
else {
render_tile(output_buf, &tile_rect);
}
});
}
}
deinit_execution();
remove_buffers_and_restore_original_inputs(orig_input_links);
}
void NodeOperation::render_tile(MemoryBuffer *output_buf, rcti *tile_rect)
{
const bool is_complex = get_flags().complex;
void *tile_data = is_complex ? initialize_tile_data(tile_rect) : nullptr;
const int elem_stride = output_buf->elem_stride;
for (int y = tile_rect->ymin; y < tile_rect->ymax; y++) {
float *output_elem = output_buf->get_elem(tile_rect->xmin, y);
if (is_complex) {
for (int x = tile_rect->xmin; x < tile_rect->xmax; x++) {
read(output_elem, x, y, tile_data);
output_elem += elem_stride;
}
}
else {
for (int x = tile_rect->xmin; x < tile_rect->xmax; x++) {
read_sampled(output_elem, x, y, PixelSampler::Nearest);
output_elem += elem_stride;
}
}
}
if (tile_data) {
deinitialize_tile_data(tile_rect, tile_data);
}
}
Vector<NodeOperationOutput *> NodeOperation::replace_inputs_with_buffers(
Span<MemoryBuffer *> inputs_bufs)
{
BLI_assert(inputs_bufs.size() == get_number_of_input_sockets());
Vector<NodeOperationOutput *> orig_links(inputs_bufs.size());
for (int i = 0; i < inputs_bufs.size(); i++) {
NodeOperationInput *input_socket = get_input_socket(i);
BufferOperation *buffer_op = new BufferOperation(inputs_bufs[i],
input_socket->get_data_type());
orig_links[i] = input_socket->get_link();
input_socket->set_link(buffer_op->get_output_socket());
buffer_op->init_execution();
}
return orig_links;
}
void NodeOperation::remove_buffers_and_restore_original_inputs(
Span<NodeOperationOutput *> original_inputs_links)
{
BLI_assert(original_inputs_links.size() == get_number_of_input_sockets());
for (int i = 0; i < original_inputs_links.size(); i++) {
NodeOperation *buffer_op = get_input_operation(i);
BLI_assert(buffer_op != nullptr);
BLI_assert(typeid(*buffer_op) == typeid(BufferOperation));
buffer_op->deinit_execution();
NodeOperationInput *input_socket = get_input_socket(i);
input_socket->set_link(original_inputs_links[i]);
delete buffer_op;
}
}
/** \} */
/*****************
@ -444,15 +289,6 @@ void NodeOperationOutput::determine_canvas(const rcti &preferred_area, rcti &r_a
std::ostream &operator<<(std::ostream &os, const NodeOperationFlags &node_operation_flags)
{
if (node_operation_flags.complex) {
os << "complex,";
}
if (node_operation_flags.open_cl) {
os << "open_cl,";
}
if (node_operation_flags.single_threaded) {
os << "single_threaded,";
}
if (node_operation_flags.use_render_border) {
os << "render_border,";
}
@ -465,12 +301,6 @@ std::ostream &operator<<(std::ostream &os, const NodeOperationFlags &node_operat
if (node_operation_flags.is_set_operation) {
os << "set_operation,";
}
if (node_operation_flags.is_write_buffer_operation) {
os << "write_buffer,";
}
if (node_operation_flags.is_read_buffer_operation) {
os << "read_buffer,";
}
if (node_operation_flags.is_proxy_operation) {
os << "proxy,";
}
@ -483,9 +313,6 @@ std::ostream &operator<<(std::ostream &os, const NodeOperationFlags &node_operat
if (!node_operation_flags.use_datatype_conversion) {
os << "no_conversion,";
}
if (node_operation_flags.is_fullframe_operation) {
os << "full_frame,";
}
if (node_operation_flags.is_constant_operation) {
os << "contant_operation,";
}
@ -505,16 +332,6 @@ std::ostream &operator<<(std::ostream &os, const NodeOperation &node_operation)
os << ",name=" << node_operation.get_name();
}
os << ",flags={" << flags << "}";
if (flags.is_read_buffer_operation) {
const ReadBufferOperation *read_operation = (const ReadBufferOperation *)&node_operation;
const MemoryProxy *proxy = read_operation->get_memory_proxy();
if (proxy) {
const WriteBufferOperation *write_operation = proxy->get_write_buffer_operation();
if (write_operation) {
os << ",write=" << (NodeOperation &)*write_operation;
}
}
}
os << ")";
return os;

238
source/blender/compositor/intern/COM_NodeOperation.h
View File

@ -22,14 +22,10 @@
#include "BKE_node.hh"
#include "BKE_node_runtime.hh"
#include "clew.h"
#include "DNA_node_types.h"
namespace blender::compositor {
class OpenCLDevice;
class ReadBufferOperation;
class ExecutionSystem;
class NodeOperation;
class NodeOperationOutput;
@ -165,28 +161,6 @@ class NodeOperationOutput {
};
struct NodeOperationFlags {
/**
* Is this an complex operation.
*
* The input and output buffers of Complex operations are stored in buffers. It allows
* sequential and read/write.
*
* Complex operations are typically doing many reads to calculate the output of a single pixel.
* Mostly Filter types (Blurs, Convolution, Defocus etc) need this to be set to true.
*/
bool complex : 1;
/**
* Does this operation support OpenCL.
*/
bool open_cl : 1;
/**
* TODO: Remove this flag and #SingleThreadedOperation if tiled implementation is removed.
* Full-frame implementation doesn't need it.
*/
bool single_threaded : 1;
/**
* Does the operation needs a viewer border.
* Basically, setting border need to happen for only operations
@ -214,8 +188,6 @@ struct NodeOperationFlags {
* TODO: To be replaced by is_constant_operation flag once tiled implementation is removed.
*/
bool is_set_operation : 1;
bool is_write_buffer_operation : 1;
bool is_read_buffer_operation : 1;
bool is_proxy_operation : 1;
bool is_viewer_operation : 1;
bool is_preview_operation : 1;
@ -228,11 +200,6 @@ struct NodeOperationFlags {
*/
bool use_datatype_conversion : 1;
/**
* Has this operation fullframe implementation.
*/
bool is_fullframe_operation : 1;
/**
* Whether operation is a primitive constant operation (Color/Vector/Value).
*/
@ -246,20 +213,14 @@ struct NodeOperationFlags {
NodeOperationFlags()
{
complex = false;
single_threaded = false;
open_cl = false;
use_render_border = false;
use_viewer_border = false;
is_canvas_set = false;
is_set_operation = false;
is_read_buffer_operation = false;
is_write_buffer_operation = false;
is_proxy_operation = false;
is_viewer_operation = false;
is_preview_operation = false;
use_datatype_conversion = true;
is_fullframe_operation = false;
is_constant_operation = false;
can_be_constant = false;
}
@ -326,28 +287,12 @@ class NodeOperation {
std::function<void(rcti &canvas)> modify_determined_canvas_fn_;
/**
* \brief mutex reference for very special node initializations
* \note only use when you really know what you are doing.
* this mutex is used to share data among chunks in the same operation
* \see TonemapOperation for an example of usage
* \see NodeOperation.init_mutex initializes this mutex
* \see NodeOperation.deinit_mutex deinitializes this mutex
* \see NodeOperation.get_mutex retrieve a pointer to this mutex.
*/
ThreadMutex mutex_;
/**
* \brief reference to the editing bNodeTree, used for break and update callback
*/
const bNodeTree *btree_;
protected:
/**
* Compositor execution model.
*/
eExecutionModel execution_model_;
rcti canvas_ = COM_AREA_NONE;
/**
@ -441,11 +386,6 @@ class NodeOperation {
return false;
}
void set_execution_model(const eExecutionModel model)
{
execution_model_ = model;
}
void set_bnodetree(const bNodeTree *tree)
{
btree_ = tree;
@ -464,58 +404,6 @@ class NodeOperation {
virtual void init_execution();
/**
* \brief when a chunk is executed by a CPUDevice, this method is called
* \ingroup execution
* \param rect: the rectangle of the chunk (location and size)
* \param chunk_number: the chunk_number to be calculated
* \param memory_buffers: all input MemoryBuffer's needed
*/
virtual void execute_region(rcti * /*rect*/, unsigned int /*chunk_number*/) {}
/**
* \brief when a chunk is executed by an OpenCLDevice, this method is called
* \ingroup execution
* \note this method is only implemented in WriteBufferOperation
* \param context: the OpenCL context
* \param program: the OpenCL program containing all compositor kernels
* \param queue: the OpenCL command queue of the device the chunk is executed on
* \param rect: the rectangle of the chunk (location and size)
* \param chunk_number: the chunk_number to be calculated
* \param memory_buffers: all input MemoryBuffer's needed
* \param output_buffer: the output-buffer to write to
*/
virtual void execute_opencl_region(OpenCLDevice * /*device*/,
rcti * /*rect*/,
unsigned int /*chunk_number*/,
MemoryBuffer ** /*memory_buffers*/,
MemoryBuffer * /*output_buffer*/)
{
}
/**
* \brief custom handle to add new tasks to the OpenCL command queue
* in order to execute a chunk on an GPUDevice.
* \ingroup execution
* \param context: the OpenCL context
* \param program: the OpenCL program containing all compositor kernels
* \param queue: the OpenCL command queue of the device the chunk is executed on
* \param output_memory_buffer: the allocated memory buffer in main CPU memory
* \param cl_output_buffer: the allocated memory buffer in OpenCLDevice memory
* \param input_memory_buffers: all input MemoryBuffer's needed
* \param cl_mem_to_clean_up: all created cl_mem references must be added to this list.
* Framework will clean this after execution
* \param cl_kernels_to_clean_up: all created cl_kernel references must be added to this list.
* Framework will clean this after execution
*/
virtual void execute_opencl(OpenCLDevice * /*device*/,
MemoryBuffer * /*output_memory_buffer*/,
cl_mem /*cl_output_buffer*/,
MemoryBuffer ** /*input_memory_buffers*/,
std::list<cl_mem> * /*cl_mem_to_clean_up*/,
std::list<cl_kernel> * /*cl_kernels_to_clean_up*/)
{
}
virtual void deinit_execution();
void set_canvas(const rcti &canvas_area);
@ -538,10 +426,6 @@ class NodeOperation {
return false;
}
virtual bool determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output);
/**
* \brief set the index of the input socket that will determine the canvas of this
* operation \param index: the index to set
@ -589,36 +473,6 @@ class NodeOperation {
return BLI_rcti_size_y(&get_canvas());
}
inline void read_sampled(float result[4], float x, float y, PixelSampler sampler)
{
execute_pixel_sampled(result, x, y, sampler);
}
inline void read_filtered(float result[4], float x, float y, float dx[2], float dy[2])
{
execute_pixel_filtered(result, x, y, dx, dy);
}
inline void read(float result[4], int x, int y, void *chunk_data)
{
execute_pixel(result, x, y, chunk_data);
}
inline void read_clamped(float result[4], int x, int y, void *chunk_data)
{
execute_pixel(result,
math::clamp(x, 0, int(this->get_width()) - 1),
math::clamp(y, 0, int(this->get_height()) - 1),
chunk_data);
}
virtual void *initialize_tile_data(rcti * /*rect*/)
{
return 0;
}
virtual void deinitialize_tile_data(rcti * /*rect*/, void * /*data*/) {}
virtual MemoryBuffer *get_input_memory_buffer(MemoryBuffer ** /*memory_buffers*/)
{
return 0;
@ -704,85 +558,9 @@ class NodeOperation {
void add_input_socket(DataType datatype, ResizeMode resize_mode = ResizeMode::Center);
void add_output_socket(DataType datatype);
/* TODO(manzanilla): to be removed with tiled implementation. */
void set_width(unsigned int width)
{
canvas_.xmax = canvas_.xmin + width;
flags_.is_canvas_set = true;
}
void set_height(unsigned int height)
{
canvas_.ymax = canvas_.ymin + height;
flags_.is_canvas_set = true;
}
SocketReader *get_input_socket_reader(unsigned int index);
void deinit_mutex();
void init_mutex();
void lock_mutex();
void unlock_mutex();
/**
* \brief set whether this operation is complex
*
* Complex operations are typically doing many reads to calculate the output of a single pixel.
* Mostly Filter types (Blurs, Convolution, Defocus etc) need this to be set to true.
*/
void set_complex(bool complex)
{
flags_.complex = complex;
}
/**
* \brief calculate a single pixel
* \note this method is called for non-complex
* \param result: is a float[4] array to store the result
* \param x: the x-coordinate of the pixel to calculate in image space
* \param y: the y-coordinate of the pixel to calculate in image space
* \param input_buffers: chunks that can be read by their ReadBufferOperation.
*/
virtual void execute_pixel_sampled(float /*output*/[4],
float /*x*/,
float /*y*/,
PixelSampler /*sampler*/)
{
}
/**
* \brief calculate a single pixel
* \note this method is called for complex
* \param result: is a float[4] array to store the result
* \param x: the x-coordinate of the pixel to calculate in image space
* \param y: the y-coordinate of the pixel to calculate in image space
* \param input_buffers: chunks that can be read by their ReadBufferOperation.
* \param chunk_data: chunk specific data a during execution time.
*/
virtual void execute_pixel(float output[4], int x, int y, void * /*chunk_data*/)
{
execute_pixel_sampled(output, x, y, PixelSampler::Nearest);
}
/**
* \brief calculate a single pixel using an EWA filter
* \note this method is called for complex
* \param result: is a float[4] array to store the result
* \param x: the x-coordinate of the pixel to calculate in image space
* \param y: the y-coordinate of the pixel to calculate in image space
* \param dx:
* \param dy:
* \param input_buffers: chunks that can be read by their ReadBufferOperation.
*/
virtual void execute_pixel_filtered(
float /*output*/[4], float /*x*/, float /*y*/, float /*dx*/[2], float /*dy*/[2])
{
}
private:
/* -------------------------------------------------------------------- */
/** \name Full Frame Methods
* \{ */
/**
* Renders given areas using operations full frame implementation.
*/
@ -790,22 +568,6 @@ class NodeOperation {
Span<rcti> areas,
Span<MemoryBuffer *> inputs_bufs);
/**
* Renders given areas using operations tiled implementation.
*/
void render_full_frame_fallback(MemoryBuffer *output_buf,
Span<rcti> areas,
Span<MemoryBuffer *> inputs);
void render_tile(MemoryBuffer *output_buf, rcti *tile_rect);
/**
* \return Replaced inputs links.
*/
Vector<NodeOperationOutput *> replace_inputs_with_buffers(Span<MemoryBuffer *> inputs_bufs);
void remove_buffers_and_restore_original_inputs(
Span<NodeOperationOutput *> original_inputs_links);
/** \} */
/* allow the DebugInfo class to look at internals */
friend class DebugInfo;

221
source/blender/compositor/intern/COM_NodeOperationBuilder.cc
View File

@ -11,14 +11,11 @@
#include "COM_Converter.h"
#include "COM_Debug.h"
#include "COM_ExecutionGroup.h"
#include "COM_PreviewOperation.h"
#include "COM_ReadBufferOperation.h"
#include "COM_SetColorOperation.h"
#include "COM_SetValueOperation.h"
#include "COM_SetVectorOperation.h"
#include "COM_ViewerOperation.h"
#include "COM_WriteBufferOperation.h"
#include "COM_ConstantFolder.h"
#include "COM_NodeOperationBuilder.h" /* own include */
@ -84,22 +81,15 @@ void NodeOperationBuilder::convert_to_operations(ExecutionSystem *system)
add_datatype_conversions();
if (context_->get_execution_model() == eExecutionModel::FullFrame) {
save_graphviz("compositor_prior_folding");
ConstantFolder folder(*this);
folder.fold_operations();
}
save_graphviz("compositor_prior_folding");
ConstantFolder folder(*this);
folder.fold_operations();
determine_canvases();
save_graphviz("compositor_prior_merging");
merge_equal_operations();
if (context_->get_execution_model() == eExecutionModel::Tiled) {
/* surround complex ops with read/write buffer */
add_complex_operation_buffers();
}
/* links not available from here on */
/* XXX make links_ a local variable to avoid confusion! */
links_.clear();
@ -109,13 +99,8 @@ void NodeOperationBuilder::convert_to_operations(ExecutionSystem *system)
/* ensure topological (link-based) order of nodes */
// sort_operations(); /* not needed yet. */
if (context_->get_execution_model() == eExecutionModel::Tiled) {
/* create execution groups */
group_operations();
}
/* transfer resulting operations to the system */
system->set_operations(operations_, groups_);
system->set_operations(operations_);
}
void NodeOperationBuilder::add_operation(NodeOperation *operation)
@ -126,7 +111,6 @@ void NodeOperationBuilder::add_operation(NodeOperation *operation)
operation->set_name(current_node_->get_bnode()->name);
operation->set_node_instance_key(current_node_->get_instance_key());
}
operation->set_execution_model(context_->get_execution_model());
operation->set_execution_system(exec_system_);
}
@ -517,133 +501,6 @@ Vector<NodeOperationInput *> NodeOperationBuilder::cache_output_links(
return inputs;
}
WriteBufferOperation *NodeOperationBuilder::find_attached_write_buffer_operation(
NodeOperationOutput *output) const
{
for (const Link &link : links_) {
if (link.from() == output) {
NodeOperation &op = link.to()->get_operation();
if (op.get_flags().is_write_buffer_operation) {
return (WriteBufferOperation *)(&op);
}
}
}
return nullptr;
}
void NodeOperationBuilder::add_input_buffers(NodeOperation * /*operation*/,
NodeOperationInput *input)
{
if (!input->is_connected()) {
return;
}
NodeOperationOutput *output = input->get_link();
if (output->get_operation().get_flags().is_read_buffer_operation) {
/* input is already buffered, no need to add another */
return;
}
/* this link will be replaced below */
remove_input_link(input);
/* check of other end already has write operation, otherwise add a new one */
WriteBufferOperation *writeoperation = find_attached_write_buffer_operation(output);
if (!writeoperation) {
writeoperation = new WriteBufferOperation(output->get_data_type());
writeoperation->set_bnodetree(context_->get_bnodetree());
add_operation(writeoperation);
add_link(output, writeoperation->get_input_socket(0));
writeoperation->read_resolution_from_input_socket();
}
/* add readbuffer op for the input */
ReadBufferOperation *readoperation = new ReadBufferOperation(output->get_data_type());
readoperation->set_memory_proxy(writeoperation->get_memory_proxy());
this->add_operation(readoperation);
add_link(readoperation->get_output_socket(), input);
readoperation->read_resolution_from_write_buffer();
}
void NodeOperationBuilder::add_output_buffers(NodeOperation *operation,
NodeOperationOutput *output)
{
/* cache connected sockets, so we can safely remove links first before replacing them */
Vector<NodeOperationInput *> targets = cache_output_links(output);
if (targets.is_empty()) {
return;
}
WriteBufferOperation *write_operation = nullptr;
for (NodeOperationInput *target : targets) {
/* try to find existing write buffer operation */
if (target->get_operation().get_flags().is_write_buffer_operation) {
BLI_assert(write_operation == nullptr); /* there should only be one write op connected */
write_operation = (WriteBufferOperation *)&target->get_operation();
}
else {
/* remove all links to other nodes */
remove_input_link(target);
}
}
/* if no write buffer operation exists yet, create a new one */
if (!write_operation) {
write_operation = new WriteBufferOperation(operation->get_output_socket()->get_data_type());
write_operation->set_bnodetree(context_->get_bnodetree());
add_operation(write_operation);
add_link(output, write_operation->get_input_socket(0));
}
write_operation->read_resolution_from_input_socket();
/* add readbuffer op for every former connected input */
for (NodeOperationInput *target : targets) {
if (&target->get_operation() == write_operation) {
continue; /* skip existing write op links */
}
ReadBufferOperation *readoperation = new ReadBufferOperation(
operation->get_output_socket()->get_data_type());
readoperation->set_memory_proxy(write_operation->get_memory_proxy());
add_operation(readoperation);
add_link(readoperation->get_output_socket(), target);
readoperation->read_resolution_from_write_buffer();
}
}
void NodeOperationBuilder::add_complex_operation_buffers()
{
/* NOTE: complex ops and get cached here first, since adding operations
* will invalidate iterators over the main operations_
*/
Vector<NodeOperation *> complex_ops;
for (NodeOperation *operation : operations_) {
if (operation->get_flags().complex) {
complex_ops.append(operation);
}
}
for (NodeOperation *op : complex_ops) {
DebugInfo::operation_read_write_buffer(op);
for (int index = 0; index < op->get_number_of_input_sockets(); index++) {
add_input_buffers(op, op->get_input_socket(index));
}
for (int index = 0; index < op->get_number_of_output_sockets(); index++) {
add_output_buffers(op, op->get_output_socket(index));
}
}
}
using Tags = std::set<NodeOperation *>;
static void find_reachable_operations_recursive(Tags &reachable, NodeOperation *op)
@ -659,13 +516,6 @@ static void find_reachable_operations_recursive(Tags &reachable, NodeOperation *
find_reachable_operations_recursive(reachable, &input->get_link()->get_operation());
}
}
/* associated write-buffer operations are executed as well */
if (op->get_flags().is_read_buffer_operation) {
ReadBufferOperation *read_op = (ReadBufferOperation *)op;
MemoryProxy *memproxy = read_op->get_memory_proxy();
find_reachable_operations_recursive(reachable, memproxy->get_write_buffer_operation());
}
}
void NodeOperationBuilder::prune_operations()
@ -725,62 +575,10 @@ void NodeOperationBuilder::sort_operations()
operations_ = sorted;
}
static void add_group_operations_recursive(Tags &visited, NodeOperation *op, ExecutionGroup *group)
{
if (visited.find(op) != visited.end()) {
return;
}
visited.insert(op);
if (!group->add_operation(op)) {
return;
}
/* add all eligible input ops to the group */
for (int i = 0; i < op->get_number_of_input_sockets(); i++) {
NodeOperationInput *input = op->get_input_socket(i);
if (input->is_connected()) {
add_group_operations_recursive(visited, &input->get_link()->get_operation(), group);
}
}
}
ExecutionGroup *NodeOperationBuilder::make_group(NodeOperation *op)
{
ExecutionGroup *group = new ExecutionGroup(groups_.size());
groups_.append(group);
Tags visited;
add_group_operations_recursive(visited, op, group);
return group;
}
void NodeOperationBuilder::group_operations()
{
for (NodeOperation *op : operations_) {
if (op->is_output_operation(context_->is_rendering())) {
ExecutionGroup *group = make_group(op);
group->set_output_execution_group(true);
}
/* add new groups for associated memory proxies where needed */
if (op->get_flags().is_read_buffer_operation) {
ReadBufferOperation *read_op = (ReadBufferOperation *)op;
MemoryProxy *memproxy = read_op->get_memory_proxy();
if (memproxy->get_executor() == nullptr) {
ExecutionGroup *group = make_group(memproxy->get_write_buffer_operation());
memproxy->set_executor(group);
}
}
}
}
void NodeOperationBuilder::save_graphviz(StringRefNull name)
{
if (COM_EXPORT_GRAPHVIZ) {
exec_system_->set_operations(operations_, groups_);
exec_system_->set_operations(operations_);
DebugInfo::graphviz(exec_system_, name);
}
}
@ -800,15 +598,6 @@ std::ostream &operator<<(std::ostream &os, const NodeOperationBuilder &builder)
os << " op" << link.from()->get_operation().get_id() << " -> op"
<< link.to()->get_operation().get_id() << ";\n";
}
for (const NodeOperation *operation : builder.get_operations()) {
if (operation->get_flags().is_read_buffer_operation) {
const ReadBufferOperation &read_operation = static_cast<const ReadBufferOperation &>(
*operation);
const WriteBufferOperation &write_operation =
*read_operation.get_memory_proxy()->get_write_buffer_operation();
os << " op" << write_operation.get_id() << " -> op" << read_operation.get_id() << ";\n";
}
}
os << "}\n";
os << "# Builder end\n";

13
source/blender/compositor/intern/COM_NodeOperationBuilder.h
View File

@ -18,13 +18,11 @@ class NodeInput;
class NodeOutput;
class ExecutionSystem;
class ExecutionGroup;
class NodeOperation;
class NodeOperationInput;
class NodeOperationOutput;
class PreviewOperation;
class WriteBufferOperation;
class ViewerOperation;
class ConstantOperation;
@ -55,7 +53,6 @@ class NodeOperationBuilder {
Vector<NodeOperation *> operations_;
Vector<Link> links_;
Vector<ExecutionGroup *> groups_;
/** Maps operation inputs to node inputs */
Map<NodeOperationInput *, NodeInput *> input_map_;
@ -133,12 +130,6 @@ class NodeOperationBuilder {
/** Helper function to store connected inputs for replacement */
Vector<NodeOperationInput *> cache_output_links(NodeOperationOutput *output) const;
/** Find a connected write buffer operation to an OpOutput */
WriteBufferOperation *find_attached_write_buffer_operation(NodeOperationOutput *output) const;
/** Add read/write buffer operations around complex operations */
void add_complex_operation_buffers();
void add_input_buffers(NodeOperation *operation, NodeOperationInput *input);
void add_output_buffers(NodeOperation *operation, NodeOperationOutput *output);
/** Remove unreachable operations */
void prune_operations();
@ -146,10 +137,6 @@ class NodeOperationBuilder {
/** Sort operations by link dependencies */
void sort_operations();
/** Create execution groups */
void group_operations();
ExecutionGroup *make_group(NodeOperation *op);
private:
PreviewOperation *make_preview_operation() const;
void unlink_inputs_and_relink_outputs(NodeOperation *unlinked_op, NodeOperation *linked_op);

271
source/blender/compositor/intern/COM_OpenCLDevice.cc
View File

@ -1,271 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_OpenCLDevice.h"
#include "COM_ExecutionGroup.h"
#include "COM_ReadBufferOperation.h"
namespace blender::compositor {
enum COM_VendorID { NVIDIA = 0x10DE, AMD = 0x1002 };
const cl_image_format IMAGE_FORMAT_COLOR = {
CL_RGBA,
CL_FLOAT,
};
const cl_image_format IMAGE_FORMAT_VECTOR = {
CL_RGB,
CL_FLOAT,
};
const cl_image_format IMAGE_FORMAT_VALUE = {
CL_R,
CL_FLOAT,
};
OpenCLDevice::OpenCLDevice(cl_context context,
cl_device_id device,
cl_program program,
cl_int vendor_id)
{
device_ = device;
context_ = context;
program_ = program;
queue_ = nullptr;
vendor_id_ = vendor_id;
cl_int error;
queue_ = clCreateCommandQueue(context_, device_, 0, &error);
}
OpenCLDevice::OpenCLDevice(OpenCLDevice &&other) noexcept
: context_(other.context_),
device_(other.device_),
program_(other.program_),
queue_(other.queue_),
vendor_id_(other.vendor_id_)
{
other.queue_ = nullptr;
}
OpenCLDevice::~OpenCLDevice()
{
if (queue_) {
clReleaseCommandQueue(queue_);
}
}
void OpenCLDevice::execute(WorkPackage *work_package)
{
const uint chunk_number = work_package->chunk_number;
ExecutionGroup *execution_group = work_package->execution_group;
MemoryBuffer **input_buffers = execution_group->get_input_buffers_opencl(chunk_number);
MemoryBuffer *output_buffer = execution_group->allocate_output_buffer(work_package->rect);
execution_group->get_output_operation()->execute_opencl_region(
this, &work_package->rect, chunk_number, input_buffers, output_buffer);
delete output_buffer;
execution_group->finalize_chunk_execution(chunk_number, input_buffers);
}
cl_mem OpenCLDevice::COM_cl_attach_memory_buffer_to_kernel_parameter(
cl_kernel kernel,
int parameter_index,
int offset_index,
std::list<cl_mem> *cleanup,
MemoryBuffer **input_memory_buffers,
SocketReader *reader)
{
return COM_cl_attach_memory_buffer_to_kernel_parameter(kernel,
parameter_index,
offset_index,
cleanup,
input_memory_buffers,
(ReadBufferOperation *)reader);
}
const cl_image_format *OpenCLDevice::determine_image_format(MemoryBuffer *memory_buffer)
{
switch (memory_buffer->get_num_channels()) {
case 1:
return &IMAGE_FORMAT_VALUE;
break;
case 3:
return &IMAGE_FORMAT_VECTOR;
break;
case 4:
return &IMAGE_FORMAT_COLOR;
break;
default:
BLI_assert_msg(0, "Unsupported num_channels.");
}
return &IMAGE_FORMAT_COLOR;
}
cl_mem OpenCLDevice::COM_cl_attach_memory_buffer_to_kernel_parameter(
cl_kernel kernel,
int parameter_index,
int offset_index,
std::list<cl_mem> *cleanup,
MemoryBuffer **input_memory_buffers,
ReadBufferOperation *reader)
{
cl_int error;
MemoryBuffer *result = reader->get_input_memory_buffer(input_memory_buffers);
const cl_image_format *image_format = determine_image_format(result);
cl_mem cl_buffer = clCreateImage2D(context_,
CL_MEM_READ_ONLY | CL_MEM_COPY_HOST_PTR,
image_format,
result->get_width(),
result->get_height(),
0,
result->get_buffer(),
&error);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
if (error == CL_SUCCESS) {
cleanup->push_back(cl_buffer);
}
error = clSetKernelArg(kernel, parameter_index, sizeof(cl_mem), &cl_buffer);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
COM_cl_attach_memory_buffer_offset_to_kernel_parameter(kernel, offset_index, result);
return cl_buffer;
}
void OpenCLDevice::COM_cl_attach_memory_buffer_offset_to_kernel_parameter(
cl_kernel kernel, int offset_index, MemoryBuffer *memory_buffer)
{
if (offset_index != -1) {
cl_int error;
const rcti &rect = memory_buffer->get_rect();
cl_int2 offset = {{rect.xmin, rect.ymin}};
error = clSetKernelArg(kernel, offset_index, sizeof(cl_int2), &offset);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
}
}
void OpenCLDevice::COM_cl_attach_size_to_kernel_parameter(cl_kernel kernel,
int offset_index,
NodeOperation *operation)
{
if (offset_index != -1) {
cl_int error;
cl_int2 offset = {{(cl_int)operation->get_width(), (cl_int)operation->get_height()}};
error = clSetKernelArg(kernel, offset_index, sizeof(cl_int2), &offset);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
}
}
void OpenCLDevice::COM_cl_attach_output_memory_buffer_to_kernel_parameter(
cl_kernel kernel, int parameter_index, cl_mem cl_output_memory_buffer)
{
cl_int error;
error = clSetKernelArg(kernel, parameter_index, sizeof(cl_mem), &cl_output_memory_buffer);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
}
void OpenCLDevice::COM_cl_enqueue_range(cl_kernel kernel, MemoryBuffer *output_memory_buffer)
{
cl_int error;
const size_t size[] = {
size_t(output_memory_buffer->get_width()),
size_t(output_memory_buffer->get_height()),
};
error = clEnqueueNDRangeKernel(queue_, kernel, 2, nullptr, size, nullptr, 0, nullptr, nullptr);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
}
void OpenCLDevice::COM_cl_enqueue_range(cl_kernel kernel,
MemoryBuffer *output_memory_buffer,
int offset_index,
NodeOperation *operation)
{
cl_int error;
const int width = output_memory_buffer->get_width();
const int height = output_memory_buffer->get_height();
int offsetx;
int offsety;
int local_size = 1024;
size_t size[2];
cl_int2 offset;
if (vendor_id_ == NVIDIA) {
local_size = 32;
}
bool breaked = false;
for (offsety = 0; offsety < height && (!breaked); offsety += local_size) {
offset.s[1] = offsety;
if (offsety + local_size < height) {
size[1] = local_size;
}
else {
size[1] = height - offsety;
}
for (offsetx = 0; offsetx < width && (!breaked); offsetx += local_size) {
if (offsetx + local_size < width) {
size[0] = local_size;
}
else {
size[0] = width - offsetx;
}
offset.s[0] = offsetx;
error = clSetKernelArg(kernel, offset_index, sizeof(cl_int2), &offset);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
error = clEnqueueNDRangeKernel(
queue_, kernel, 2, nullptr, size, nullptr, 0, nullptr, nullptr);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
clFlush(queue_);
if (operation->is_braked()) {
breaked = false;
}
}
}
}
cl_kernel OpenCLDevice::COM_cl_create_kernel(const char *kernelname,
std::list<cl_kernel> *cl_kernels_to_clean_up)
{
cl_int error;
cl_kernel kernel = clCreateKernel(program_, kernelname, &error);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
else {
if (cl_kernels_to_clean_up) {
cl_kernels_to_clean_up->push_back(kernel);
}
}
return kernel;
}
} // namespace blender::compositor

120
source/blender/compositor/intern/COM_OpenCLDevice.h
View File

@ -1,120 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
class OpenCLDevice;
#pragma once
#include <list>
#include "COM_Device.h"
#include "clew.h"
namespace blender::compositor {
class NodeOperation;
class MemoryBuffer;
class ReadBufferOperation;
typedef NodeOperation SocketReader;
/**
* \brief device representing an GPU OpenCL device.
* an instance of this class represents a single cl_device
*/
class OpenCLDevice : public Device {
private:
/**
* \brief OPENCL context
*/
cl_context context_;
/**
* \brief OPENCL device
*/
cl_device_id device_;
/**
* \brief OPENCL program
*/
cl_program program_;
/**
* \brief OPENCL command queue
*/
cl_command_queue queue_;
/**
* \brief OPENCL vendor ID
*/
cl_int vendor_id_;
public:
/**
* \brief constructor with OPENCL device
* \param context:
* \param device:
* \param program:
* \param vendorID:
*/
OpenCLDevice(cl_context context, cl_device_id device, cl_program program, cl_int vendor_id);
OpenCLDevice(OpenCLDevice &&other) noexcept;
~OpenCLDevice();
/**
* \brief execute a WorkPackage
* \param work: the WorkPackage to execute
*/
void execute(WorkPackage *work) override;
/**
* \brief determine an image format
* \param memorybuffer:
*/
static const cl_image_format *determine_image_format(MemoryBuffer *memory_buffer);
cl_context get_context()
{
return context_;
}
cl_command_queue get_queue()
{
return queue_;
}
cl_mem COM_cl_attach_memory_buffer_to_kernel_parameter(cl_kernel kernel,
int parameter_index,
int offset_index,
std::list<cl_mem> *cleanup,
MemoryBuffer **input_memory_buffers,
SocketReader *reader);
cl_mem COM_cl_attach_memory_buffer_to_kernel_parameter(cl_kernel kernel,
int parameter_index,
int offset_index,
std::list<cl_mem> *cleanup,
MemoryBuffer **input_memory_buffers,
ReadBufferOperation *reader);
void COM_cl_attach_memory_buffer_offset_to_kernel_parameter(cl_kernel kernel,
int offset_index,
MemoryBuffer *memory_buffers);
void COM_cl_attach_output_memory_buffer_to_kernel_parameter(cl_kernel kernel,
int parameter_index,
cl_mem cl_output_memory_buffer);
void COM_cl_attach_size_to_kernel_parameter(cl_kernel kernel,
int offset_index,
NodeOperation *operation);
void COM_cl_enqueue_range(cl_kernel kernel, MemoryBuffer *output_memory_buffer);
void COM_cl_enqueue_range(cl_kernel kernel,
MemoryBuffer *output_memory_buffer,
int offset_index,
NodeOperation *operation);
cl_kernel COM_cl_create_kernel(const char *kernelname,
std::list<cl_kernel> *cl_kernels_to_clean_up);
};
} // namespace blender::compositor

49
source/blender/compositor/intern/COM_SingleThreadedOperation.cc
View File

@ -1,49 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_SingleThreadedOperation.h"
namespace blender::compositor {
SingleThreadedOperation::SingleThreadedOperation()
{
cached_instance_ = nullptr;
flags_.complex = true;
flags_.single_threaded = true;
}
void SingleThreadedOperation::init_execution()
{
init_mutex();
}
void SingleThreadedOperation::execute_pixel(float output[4], int x, int y, void * /*data*/)
{
cached_instance_->read_no_check(output, x, y);
}
void SingleThreadedOperation::deinit_execution()
{
deinit_mutex();
if (cached_instance_) {
delete cached_instance_;
cached_instance_ = nullptr;
}
}
void *SingleThreadedOperation::initialize_tile_data(rcti *rect)
{
if (cached_instance_) {
return cached_instance_;
}
lock_mutex();
if (cached_instance_ == nullptr) {
//
cached_instance_ = create_memory_buffer(rect);
}
unlock_mutex();
return cached_instance_;
}
} // namespace blender::compositor

44
source/blender/compositor/intern/COM_SingleThreadedOperation.h
View File

@ -1,44 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#include "COM_NodeOperation.h"
namespace blender::compositor {
class SingleThreadedOperation : public NodeOperation {
private:
MemoryBuffer *cached_instance_;
protected:
inline bool is_cached()
{
return cached_instance_ != nullptr;
}
public:
SingleThreadedOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel(float output[4], int x, int y, void *data) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void *initialize_tile_data(rcti *rect) override;
virtual MemoryBuffer *create_memory_buffer(rcti *rect) = 0;
};
} // namespace blender::compositor

148
source/blender/compositor/intern/COM_TiledExecutionModel.cc
View File

@ -1,148 +0,0 @@
/* SPDX-FileCopyrightText: 2021 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_TiledExecutionModel.h"
#include "COM_Debug.h"
#include "COM_ExecutionGroup.h"
#include "COM_ReadBufferOperation.h"
#include "COM_WorkScheduler.h"
#include "BLT_translation.hh"
#ifdef WITH_CXX_GUARDEDALLOC
# include "MEM_guardedalloc.h"
#endif
namespace blender::compositor {
TiledExecutionModel::TiledExecutionModel(CompositorContext &context,
Span<NodeOperation *> operations,
Span<ExecutionGroup *> groups)
: ExecutionModel(context, operations), groups_(groups)
{
const bNodeTree *node_tree = context.get_bnodetree();
node_tree->runtime->stats_draw(node_tree->runtime->sdh,
RPT_("Compositing | Determining resolution"));
uint resolution[2];
for (ExecutionGroup *group : groups_) {
resolution[0] = 0;
resolution[1] = 0;
group->determine_resolution(resolution);
if (border_.use_render_border) {
const rctf *render_border = border_.render_border;
group->set_render_border(
render_border->xmin, render_border->xmax, render_border->ymin, render_border->ymax);
}
if (border_.use_viewer_border) {
const rctf *viewer_border = border_.viewer_border;
group->set_viewer_border(
viewer_border->xmin, viewer_border->xmax, viewer_border->ymin, viewer_border->ymax);
}
}
}
static void update_read_buffer_offset(Span<NodeOperation *> operations)
{
uint order = 0;
for (NodeOperation *operation : operations) {
if (operation->get_flags().is_read_buffer_operation) {
ReadBufferOperation *read_operation = (ReadBufferOperation *)operation;
read_operation->set_offset(order);
order++;
}
}
}
static void init_write_operations_for_execution(Span<NodeOperation *> operations,
const bNodeTree *bTree)
{
for (NodeOperation *operation : operations) {
if (operation->get_flags().is_write_buffer_operation) {
operation->set_bnodetree(bTree);
operation->init_execution();
}
}
}
static void link_write_buffers(Span<NodeOperation *> operations)
{
for (NodeOperation *operation : operations) {
if (operation->get_flags().is_read_buffer_operation) {
ReadBufferOperation *read_operation = static_cast<ReadBufferOperation *>(operation);
read_operation->update_memory_buffer();
}
}
}
static void init_non_write_operations_for_execution(Span<NodeOperation *> operations,
const bNodeTree *bTree)
{
for (NodeOperation *operation : operations) {
if (!operation->get_flags().is_write_buffer_operation) {
operation->set_bnodetree(bTree);
operation->init_execution();
}
}
}
static void init_execution_groups_for_execution(Span<ExecutionGroup *> groups,
const int chunk_size)
{
for (ExecutionGroup *execution_group : groups) {
execution_group->set_chunksize(chunk_size);
execution_group->init_execution();
}
}
void TiledExecutionModel::execute(ExecutionSystem &exec_system)
{
const bNodeTree *editingtree = this->context_.get_bnodetree();
editingtree->runtime->stats_draw(editingtree->runtime->sdh,
RPT_("Compositing | Initializing execution"));
update_read_buffer_offset(operations_);
init_write_operations_for_execution(operations_, context_.get_bnodetree());
link_write_buffers(operations_);
init_non_write_operations_for_execution(operations_, context_.get_bnodetree());
init_execution_groups_for_execution(groups_, context_.get_chunksize());
WorkScheduler::start(context_);
execute_groups(eCompositorPriority::High, exec_system);
if (!context_.is_fast_calculation()) {
execute_groups(eCompositorPriority::Medium, exec_system);
execute_groups(eCompositorPriority::Low, exec_system);
}
WorkScheduler::finish();
WorkScheduler::stop();
editingtree->runtime->stats_draw(editingtree->runtime->sdh,
RPT_("Compositing | De-initializing execution"));
for (NodeOperation *operation : operations_) {
operation->deinit_execution();
}
for (ExecutionGroup *execution_group : groups_) {
execution_group->deinit_execution();
}
}
void TiledExecutionModel::execute_groups(eCompositorPriority priority,
ExecutionSystem &exec_system)
{
for (ExecutionGroup *execution_group : groups_) {
if (execution_group->get_flags().is_output &&
execution_group->get_render_priority() == priority)
{
execution_group->execute(&exec_system);
}
}
}
} // namespace blender::compositor

41
source/blender/compositor/intern/COM_TiledExecutionModel.h
View File

@ -1,41 +0,0 @@
/* SPDX-FileCopyrightText: 2021 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#pragma once
#include "COM_Enums.h"
#include "COM_ExecutionModel.h"
#ifdef WITH_CXX_GUARDEDALLOC
# include "MEM_guardedalloc.h"
#endif
namespace blender::compositor {
class ExecutionGroup;
/**
* Operations are executed from outputs to inputs grouped in execution groups and rendered in
* tiles.
*/
class TiledExecutionModel : public ExecutionModel {
private:
Span<ExecutionGroup *> groups_;
public:
TiledExecutionModel(CompositorContext &context,
Span<NodeOperation *> operations,
Span<ExecutionGroup *> groups);
void execute(ExecutionSystem &exec_system) override;
private:
void execute_groups(eCompositorPriority priority, ExecutionSystem &exec_system);
#ifdef WITH_CXX_GUARDEDALLOC
MEM_CXX_CLASS_ALLOC_FUNCS("COM:TiledExecutionModel")
#endif
};
} // namespace blender::compositor

22
source/blender/compositor/intern/COM_WorkPackage.cc
View File

@ -1,22 +0,0 @@
/* SPDX-FileCopyrightText: 2011 Blender Authors
*
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_WorkPackage.h"
#include "COM_ExecutionGroup.h"
namespace blender::compositor {
std::ostream &operator<<(std::ostream &os, const WorkPackage &work_package)
{
os << "WorkPackage(execution_group=" << *work_package.execution_group;
os << ",chunk=" << work_package.chunk_number;
os << ",state=" << work_package.state;
os << ",rect=(" << work_package.rect.xmin << "," << work_package.rect.ymin << ")-("
<< work_package.rect.xmax << "," << work_package.rect.ymax << ")";
os << ")";
return os;
}
} // namespace blender::compositor

30
source/blender/compositor/intern/COM_WorkPackage.h
View File

@ -8,43 +8,17 @@
# include "MEM_guardedalloc.h"
#endif
#include "COM_Enums.h"
#include "DNA_vec_types.h"
#include <functional>
#include <ostream>
namespace blender::compositor {
/* Forward Declarations. */
class ExecutionGroup;
/**
* \brief contains data about work that can be scheduled
* \see WorkScheduler
*/
struct WorkPackage {
eWorkPackageType type;
eWorkPackageState state = eWorkPackageState::NotScheduled;
/**
* \brief execution_group with the operations-setup to be evaluated
*/
ExecutionGroup *execution_group;
/**
* \brief number of the chunk to be executed
*/
unsigned int chunk_number;
/**
* Area of the execution group that the work package calculates.
*/
rcti rect;
/**
* Custom function to execute when work package type is CustomFunction.
* Called to execute work.
*/
std::function<void()> execute_fn;
@ -58,6 +32,4 @@ struct WorkPackage {
#endif
};
std::ostream &operator<<(std::ostream &os, const WorkPackage &work_package);
} // namespace blender::compositor

266
source/blender/compositor/intern/COM_WorkScheduler.cc
View File

@ -5,13 +5,6 @@
#include "COM_WorkScheduler.h"
#include "COM_CPUDevice.h"
#include "COM_CompositorContext.h"
#include "COM_ExecutionGroup.h"
#include "COM_OpenCLDevice.h"
#include "COM_OpenCLKernels.cl.h"
#include "COM_WriteBufferOperation.h"
#include "clew.h"
#include "MEM_guardedalloc.h"
@ -42,14 +35,6 @@ constexpr ThreadingModel COM_threading_model()
return ThreadingModel::Queue;
}
/**
* Does the active threading model support opencl?
*/
constexpr bool COM_is_opencl_enabled()
{
return COM_threading_model() != ThreadingModel::SingleThreaded;
}
static ThreadLocal(CPUDevice *) g_thread_device;
static struct {
struct {
@ -69,222 +54,9 @@ static struct {
TaskPool *pool;
} task;
struct {
ThreadQueue *queue;
cl_context context;
cl_program program;
/** \brief list of all OpenCLDevices. for every OpenCL GPU device an instance of OpenCLDevice
* is created. */
Vector<OpenCLDevice> devices;
/** \brief list of all thread for every GPUDevice in cpudevices a thread exists. */
ListBase threads;
/** \brief all scheduled work for the GPU. */
bool active = false;
bool initialized = false;
} opencl;
int num_cpu_threads;
} g_work_scheduler;
/* -------------------------------------------------------------------- */
/** \name OpenCL Scheduling
* \{ */
static void CL_CALLBACK cl_context_error(const char *errinfo,
const void * /*private_info*/,
size_t /*cb*/,
void * /*user_data*/)
{
printf("OPENCL error: %s\n", errinfo);
}
static void *thread_execute_gpu(void *data)
{
Device *device = (Device *)data;
WorkPackage *work;
while ((work = (WorkPackage *)BLI_thread_queue_pop(g_work_scheduler.opencl.queue))) {
device->execute(work);
}
return nullptr;
}
static void opencl_start(const CompositorContext &context)
{
if (context.get_has_active_opencl_devices()) {
g_work_scheduler.opencl.queue = BLI_thread_queue_init();
BLI_threadpool_init(&g_work_scheduler.opencl.threads,
thread_execute_gpu,
g_work_scheduler.opencl.devices.size());
for (Device &device : g_work_scheduler.opencl.devices) {
BLI_threadpool_insert(&g_work_scheduler.opencl.threads, &device);
}
g_work_scheduler.opencl.active = true;
}
else {
g_work_scheduler.opencl.active = false;
}
}
static bool opencl_schedule(WorkPackage *package)
{
if (package->type == eWorkPackageType::Tile && package->execution_group->get_flags().open_cl &&
g_work_scheduler.opencl.active)
{
BLI_thread_queue_push(g_work_scheduler.opencl.queue, package);
return true;
}
return false;
}
static void opencl_finish()
{
if (g_work_scheduler.opencl.active) {
BLI_thread_queue_wait_finish(g_work_scheduler.opencl.queue);
}
}
static void opencl_stop()
{
if (g_work_scheduler.opencl.active) {
BLI_thread_queue_nowait(g_work_scheduler.opencl.queue);
BLI_threadpool_end(&g_work_scheduler.opencl.threads);
BLI_thread_queue_free(g_work_scheduler.opencl.queue);
g_work_scheduler.opencl.queue = nullptr;
}
}
static bool opencl_has_gpu_devices()
{
return !g_work_scheduler.opencl.devices.is_empty();
}
static void opencl_initialize(const bool use_opencl)
{
/* deinitialize OpenCL GPU's */
if (use_opencl && !g_work_scheduler.opencl.initialized) {
g_work_scheduler.opencl.context = nullptr;
g_work_scheduler.opencl.program = nullptr;
/* This will check for errors and skip if already initialized. */
if (clewInit() != CLEW_SUCCESS) {
return;
}
if (clCreateContextFromType) {
cl_uint number_of_platforms = 0;
cl_int error;
error = clGetPlatformIDs(0, nullptr, &number_of_platforms);
if (error == -1001) {
} /* GPU not supported */
else if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
if (G.f & G_DEBUG) {
printf("%u number of platforms\n", number_of_platforms);
}
cl_platform_id *platforms = (cl_platform_id *)MEM_mallocN(
sizeof(cl_platform_id) * number_of_platforms, __func__);
error = clGetPlatformIDs(number_of_platforms, platforms, nullptr);
uint index_platform;
for (index_platform = 0; index_platform < number_of_platforms; index_platform++) {
cl_platform_id platform = platforms[index_platform];
cl_uint number_of_devices = 0;
clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, 0, nullptr, &number_of_devices);
if (number_of_devices <= 0) {
continue;
}
cl_device_id *cldevices = (cl_device_id *)MEM_mallocN(
sizeof(cl_device_id) * number_of_devices, __func__);
clGetDeviceIDs(platform, CL_DEVICE_TYPE_GPU, number_of_devices, cldevices, nullptr);
g_work_scheduler.opencl.context = clCreateContext(
nullptr, number_of_devices, cldevices, cl_context_error, nullptr, &error);
if (error != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
const char *cl_str[2] = {datatoc_COM_OpenCLKernels_cl, nullptr};
g_work_scheduler.opencl.program = clCreateProgramWithSource(
g_work_scheduler.opencl.context, 1, cl_str, nullptr, &error);
error = clBuildProgram(g_work_scheduler.opencl.program,
number_of_devices,
cldevices,
nullptr,
nullptr,
nullptr);
if (error != CL_SUCCESS) {
cl_int error2;
size_t ret_val_size = 0;
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
error2 = clGetProgramBuildInfo(g_work_scheduler.opencl.program,
cldevices[0],
CL_PROGRAM_BUILD_LOG,
0,
nullptr,
&ret_val_size);
if (error2 != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
char *build_log = (char *)MEM_mallocN(sizeof(char) * ret_val_size + 1, __func__);
error2 = clGetProgramBuildInfo(g_work_scheduler.opencl.program,
cldevices[0],
CL_PROGRAM_BUILD_LOG,
ret_val_size,
build_log,
nullptr);
if (error2 != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error, clewErrorString(error));
}
build_log[ret_val_size] = '\0';
printf("%s", build_log);
MEM_freeN(build_log);
}
else {
uint index_devices;
for (index_devices = 0; index_devices < number_of_devices; index_devices++) {
cl_device_id device = cldevices[index_devices];
cl_int vendorID = 0;
cl_int error2 = clGetDeviceInfo(
device, CL_DEVICE_VENDOR_ID, sizeof(cl_int), &vendorID, nullptr);
if (error2 != CL_SUCCESS) {
printf("CLERROR[%d]: %s\n", error2, clewErrorString(error2));
}
g_work_scheduler.opencl.devices.append_as(g_work_scheduler.opencl.context,
device,
g_work_scheduler.opencl.program,
vendorID);
}
}
MEM_freeN(cldevices);
}
MEM_freeN(platforms);
}
g_work_scheduler.opencl.initialized = true;
}
}
static void opencl_deinitialize()
{
g_work_scheduler.opencl.devices.clear_and_shrink();
if (g_work_scheduler.opencl.program) {
clReleaseProgram(g_work_scheduler.opencl.program);
g_work_scheduler.opencl.program = nullptr;
}
if (g_work_scheduler.opencl.context) {
clReleaseContext(g_work_scheduler.opencl.context);
g_work_scheduler.opencl.context = nullptr;
}
g_work_scheduler.opencl.initialized = false;
}
/** \} */
/* -------------------------------------------------------------------- */
/** \name Single threaded Scheduling
* \{ */
@ -419,12 +191,6 @@ static void threading_model_task_stop()
void WorkScheduler::schedule(WorkPackage *package)
{
if (COM_is_opencl_enabled()) {
if (opencl_schedule(package)) {
return;
}
}
switch (COM_threading_model()) {
case ThreadingModel::SingleThreaded: {
threading_model_single_thread_execute(package);
@ -443,12 +209,8 @@ void WorkScheduler::schedule(WorkPackage *package)
}
}
void WorkScheduler::start(const CompositorContext &context)
void WorkScheduler::start()
{
if (COM_is_opencl_enabled()) {
opencl_start(context);
}
switch (COM_threading_model()) {
case ThreadingModel::SingleThreaded:
/* Nothing to do. */
@ -466,10 +228,6 @@ void WorkScheduler::start(const CompositorContext &context)
void WorkScheduler::finish()
{
if (COM_is_opencl_enabled()) {
opencl_finish();
}
switch (COM_threading_model()) {
case ThreadingModel::SingleThreaded:
/* Nothing to do. */
@ -487,10 +245,6 @@ void WorkScheduler::finish()
void WorkScheduler::stop()
{
if (COM_is_opencl_enabled()) {
opencl_stop();
}
switch (COM_threading_model()) {
case ThreadingModel::SingleThreaded:
/* Nothing to do. */
@ -506,20 +260,8 @@ void WorkScheduler::stop()
}
}
bool WorkScheduler::has_gpu_devices()
void WorkScheduler::initialize(int num_cpu_threads)
{
if (COM_is_opencl_enabled()) {
return opencl_has_gpu_devices();
}
return false;
}
void WorkScheduler::initialize(bool use_opencl, int num_cpu_threads)
{
if (COM_is_opencl_enabled()) {
opencl_initialize(use_opencl);
}
g_work_scheduler.num_cpu_threads = num_cpu_threads;
switch (COM_threading_model()) {
case ThreadingModel::SingleThreaded:
@ -538,10 +280,6 @@ void WorkScheduler::initialize(bool use_opencl, int num_cpu_threads)
void WorkScheduler::deinitialize()
{
if (COM_is_opencl_enabled()) {
opencl_deinitialize();
}
switch (COM_threading_model()) {
case ThreadingModel::SingleThreaded:
/* Nothing to do. */

23
source/blender/compositor/intern/COM_WorkScheduler.h
View File

@ -12,8 +12,6 @@ namespace blender::compositor {
struct WorkPackage;
class CompositorContext;
/** \brief the workscheduler
* \ingroup execution
*/
@ -21,9 +19,6 @@ struct WorkScheduler {
/**
* \brief schedule a chunk of a group to be calculated.
* An execution group schedules a chunk in the WorkScheduler
* when ExecutionGroup.get_flags().open_cl is set the work will be handled by a OpenCLDevice
* otherwise the work is scheduled for an CPUDevice
* \see ExecutionGroup.execute
*/
static void schedule(WorkPackage *package);
@ -33,13 +28,9 @@ struct WorkScheduler {
* during initialization the mutexes are initialized.
* there are two mutexes (for every device type one)
* After mutex initialization the system is queried in order to count the number of CPUDevices
* and GPUDevices to be created. For every hardware thread a CPUDevice and for every OpenCL GPU
* device a OpenCLDevice is created. these devices are stored in a separate list (cpudevices &
* gpudevices)
*
* This function can be called multiple times to lazily initialize OpenCL.
* to be created. For every hardware thread a CPUDevice is created.
*/
static void initialize(bool use_opencl, int num_cpu_threads);
static void initialize(int num_cpu_threads);
/**
* \brief deinitialize the WorkScheduler
@ -53,7 +44,7 @@ struct WorkScheduler {
* for every device a thread is created.
* \see initialize Initialization and query of the number of devices
*/
static void start(const CompositorContext &context);
static void start();
/**
* \brief stop the execution
@ -67,14 +58,6 @@ struct WorkScheduler {
*/
static void finish();
/**
* \brief Are there OpenCL capable GPU devices initialized?
* the result of this method is stored in the CompositorContext
* A node can generate a different operation tree when OpenCLDevices exists.
* \see CompositorContext.get_has_active_opencl_devices
*/
static bool has_gpu_devices();
static int get_num_cpu_threads();
static int current_thread_id();

2
source/blender/compositor/intern/COM_compositor.cc
View File

@ -89,7 +89,7 @@ void COM_execute(Render *render,
/* CPU compositor. */
/* Initialize workscheduler. */
blender::compositor::WorkScheduler::initialize(false, BKE_render_num_threads(render_data));
blender::compositor::WorkScheduler::initialize(BKE_render_num_threads(render_data));
/* Execute. */
const bool twopass = (node_tree->flag & NTREE_TWO_PASS) && !rendering;

8
source/blender/compositor/nodes/COM_BlurNode.cc
View File

@ -5,11 +5,9 @@
#include "COM_BlurNode.h"
#include "COM_FastGaussianBlurOperation.h"
#include "COM_GammaCorrectOperation.h"
#include "COM_GaussianAlphaXBlurOperation.h"
#include "COM_GaussianAlphaYBlurOperation.h"
#include "COM_GaussianAlphaBlurBaseOperation.h"
#include "COM_GaussianBlurBaseOperation.h"
#include "COM_GaussianBokehBlurOperation.h"
#include "COM_GaussianXBlurOperation.h"
#include "COM_GaussianYBlurOperation.h"
#include "COM_MathBaseOperation.h"
#include "COM_SetValueOperation.h"
@ -93,7 +91,6 @@ void BlurNode::convert_to_operations(NodeConverter &converter,
GaussianXBlurOperation *operationx = new GaussianXBlurOperation();
operationx->set_data(data);
operationx->set_quality(quality);
operationx->check_opencl();
operationx->set_extend_bounds(extend_bounds);
converter.add_operation(operationx);
@ -102,7 +99,6 @@ void BlurNode::convert_to_operations(NodeConverter &converter,
GaussianYBlurOperation *operationy = new GaussianYBlurOperation();
operationy->set_data(data);
operationy->set_quality(quality);
operationy->check_opencl();
operationy->set_extend_bounds(extend_bounds);
converter.add_operation(operationy);

3
source/blender/compositor/nodes/COM_DilateErodeNode.cc
View File

@ -4,8 +4,7 @@
#include "COM_DilateErodeNode.h"
#include "COM_DilateErodeOperation.h"
#include "COM_GaussianAlphaXBlurOperation.h"
#include "COM_GaussianAlphaYBlurOperation.h"
#include "COM_GaussianAlphaBlurBaseOperation.h"
#include "COM_SMAAOperation.h"
namespace blender::compositor {

3
source/blender/compositor/nodes/COM_KeyingNode.cc
View File

@ -18,8 +18,7 @@
#include "COM_SetAlphaMultiplyOperation.h"
#include "COM_GaussianAlphaXBlurOperation.h"
#include "COM_GaussianAlphaYBlurOperation.h"
#include "COM_GaussianAlphaBlurBaseOperation.h"
#include "BLI_math_color.h"

3
source/blender/compositor/nodes/COM_KuwaharaNode.cc
View File

@ -9,8 +9,7 @@
#include "COM_KuwaharaNode.h"
#include "COM_GaussianXBlurOperation.h"
#include "COM_GaussianYBlurOperation.h"
#include "COM_GaussianBlurBaseOperation.h"
#include "COM_KuwaharaAnisotropicOperation.h"
#include "COM_KuwaharaAnisotropicStructureTensorOperation.h"
#include "COM_KuwaharaClassicOperation.h"

19
source/blender/compositor/nodes/COM_RotateNode.cc
View File

@ -15,7 +15,7 @@ RotateNode::RotateNode(bNode *editor_node) : Node(editor_node)
}
void RotateNode::convert_to_operations(NodeConverter &converter,
const CompositorContext &context) const
const CompositorContext & /*context*/) const
{
NodeInput *input_socket = this->get_input_socket(0);
NodeInput *input_degree_socket = this->get_input_socket(1);
@ -24,21 +24,8 @@ void RotateNode::convert_to_operations(NodeConverter &converter,
converter.add_operation(operation);
PixelSampler sampler = (PixelSampler)this->get_bnode()->custom1;
switch (context.get_execution_model()) {
case eExecutionModel::Tiled: {
SetSamplerOperation *sampler_op = new SetSamplerOperation();
sampler_op->set_sampler(sampler);
converter.add_operation(sampler_op);
converter.add_link(sampler_op->get_output_socket(), operation->get_input_socket(0));
converter.map_input_socket(input_socket, sampler_op->get_input_socket(0));
break;
}
case eExecutionModel::FullFrame: {
operation->set_sampler(sampler);
converter.map_input_socket(input_socket, operation->get_input_socket(0));
break;
}
}
operation->set_sampler(sampler);
converter.map_input_socket(input_socket, operation->get_input_socket(0));
converter.map_input_socket(input_degree_socket, operation->get_input_socket(1));
converter.map_output_socket(output_socket, operation->get_output_socket(0));

45
source/blender/compositor/nodes/COM_SocketProxyNode.cc
View File

@ -3,8 +3,6 @@
* SPDX-License-Identifier: GPL-2.0-or-later */
#include "COM_SocketProxyNode.h"
#include "COM_ReadBufferOperation.h"
#include "COM_WriteBufferOperation.h"
namespace blender::compositor {
@ -43,47 +41,4 @@ void SocketProxyNode::convert_to_operations(NodeConverter &converter,
converter.map_output_socket(get_output_socket(), proxy_output);
}
SocketBufferNode::SocketBufferNode(bNode *editor_node,
bNodeSocket *editor_input,
bNodeSocket *editor_output)
: Node(editor_node, false)
{
DataType dt;
dt = DataType::Value;
if (editor_input->type == SOCK_RGBA) {
dt = DataType::Color;
}
if (editor_input->type == SOCK_VECTOR) {
dt = DataType::Vector;
}
this->add_input_socket(dt, editor_input);
dt = DataType::Value;
if (editor_output->type == SOCK_RGBA) {
dt = DataType::Color;
}
if (editor_output->type == SOCK_VECTOR) {
dt = DataType::Vector;
}
this->add_output_socket(dt, editor_output);
}
void SocketBufferNode::convert_to_operations(NodeConverter &converter,
const CompositorContext & /*context*/) const
{
NodeOutput *output = this->get_output_socket(0);
NodeInput *input = this->get_input_socket(0);
DataType datatype = output->get_data_type();
WriteBufferOperation *write_operation = new WriteBufferOperation(datatype);
ReadBufferOperation *read_operation = new ReadBufferOperation(datatype);
read_operation->set_memory_proxy(write_operation->get_memory_proxy());
converter.add_operation(write_operation);
converter.add_operation(read_operation);
converter.map_input_socket(input, write_operation->get_input_socket(0));
converter.map_output_socket(output, read_operation->get_output_socket());
}
} // namespace blender::compositor

7
source/blender/compositor/nodes/COM_SocketProxyNode.h
View File

@ -35,11 +35,4 @@ class SocketProxyNode : public Node {
bool use_conversion_;
};
class SocketBufferNode : public Node {
public:
SocketBufferNode(bNode *editor_node, bNodeSocket *editor_input, bNodeSocket *editor_output);
void convert_to_operations(NodeConverter &converter,
const CompositorContext &context) const override;
};
} // namespace blender::compositor

146
source/blender/compositor/nodes/COM_Stabilize2dNode.cc
View File

@ -55,119 +55,55 @@ void Stabilize2dNode::convert_to_operations(NodeConverter &converter,
converter.add_operation(x_attribute);
converter.add_operation(y_attribute);
switch (context.get_execution_model()) {
case eExecutionModel::Tiled: {
ScaleRelativeOperation *scale_operation = new ScaleRelativeOperation();
scale_operation->set_sampler(sampler);
RotateOperation *rotate_operation = new RotateOperation();
rotate_operation->set_do_degree2_rad_conversion(false);
TranslateOperation *translate_operation = new TranslateOperation();
SetSamplerOperation *psoperation = new SetSamplerOperation();
psoperation->set_sampler(sampler);
ScaleRelativeOperation *scale_operation = new ScaleRelativeOperation();
scale_operation->set_sampler(sampler);
RotateOperation *rotate_operation = new RotateOperation();
rotate_operation->set_do_degree2_rad_conversion(false);
rotate_operation->set_sampler(sampler);
TranslateOperation *translate_operation = new TranslateCanvasOperation();
converter.add_operation(scale_operation);
converter.add_operation(translate_operation);
converter.add_operation(rotate_operation);
converter.add_operation(psoperation);
converter.add_operation(scale_operation);
converter.add_operation(translate_operation);
converter.add_operation(rotate_operation);
converter.add_link(scale_attribute->get_output_socket(),
scale_operation->get_input_socket(1));
converter.add_link(scale_attribute->get_output_socket(),
scale_operation->get_input_socket(2));
converter.add_link(scale_attribute->get_output_socket(), scale_operation->get_input_socket(1));
converter.add_link(scale_attribute->get_output_socket(), scale_operation->get_input_socket(2));
converter.add_link(angle_attribute->get_output_socket(),
rotate_operation->get_input_socket(1));
converter.add_link(angle_attribute->get_output_socket(), rotate_operation->get_input_socket(1));
converter.add_link(x_attribute->get_output_socket(),
translate_operation->get_input_socket(1));
converter.add_link(y_attribute->get_output_socket(),
translate_operation->get_input_socket(2));
converter.add_link(x_attribute->get_output_socket(), translate_operation->get_input_socket(1));
converter.add_link(y_attribute->get_output_socket(), translate_operation->get_input_socket(2));
converter.map_output_socket(get_output_socket(), psoperation->get_output_socket());
NodeOperationInput *stabilization_socket = nullptr;
if (invert) {
/* Translate -> Rotate -> Scale. */
stabilization_socket = translate_operation->get_input_socket(0);
converter.map_input_socket(image_input, translate_operation->get_input_socket(0));
if (invert) {
/* Translate -> Rotate -> Scale. */
converter.map_input_socket(image_input, translate_operation->get_input_socket(0));
converter.add_link(translate_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.add_link(rotate_operation->get_output_socket(),
scale_operation->get_input_socket(0));
converter.add_link(translate_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.add_link(rotate_operation->get_output_socket(),
scale_operation->get_input_socket(0));
converter.add_link(scale_operation->get_output_socket(), psoperation->get_input_socket(0));
}
else {
/* Scale -> Rotate -> Translate. */
converter.map_input_socket(image_input, scale_operation->get_input_socket(0));
converter.add_link(scale_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.add_link(rotate_operation->get_output_socket(),
translate_operation->get_input_socket(0));
converter.add_link(translate_operation->get_output_socket(),
psoperation->get_input_socket(0));
}
break;
}
case eExecutionModel::FullFrame: {
ScaleRelativeOperation *scale_operation = new ScaleRelativeOperation();
scale_operation->set_sampler(sampler);
RotateOperation *rotate_operation = new RotateOperation();
rotate_operation->set_do_degree2_rad_conversion(false);
rotate_operation->set_sampler(sampler);
TranslateOperation *translate_operation = new TranslateCanvasOperation();
converter.add_operation(scale_operation);
converter.add_operation(translate_operation);
converter.add_operation(rotate_operation);
converter.add_link(scale_attribute->get_output_socket(),
scale_operation->get_input_socket(1));
converter.add_link(scale_attribute->get_output_socket(),
scale_operation->get_input_socket(2));
converter.add_link(angle_attribute->get_output_socket(),
rotate_operation->get_input_socket(1));
converter.add_link(x_attribute->get_output_socket(),
translate_operation->get_input_socket(1));
converter.add_link(y_attribute->get_output_socket(),
translate_operation->get_input_socket(2));
NodeOperationInput *stabilization_socket = nullptr;
if (invert) {
/* Translate -> Rotate -> Scale. */
stabilization_socket = translate_operation->get_input_socket(0);
converter.map_input_socket(image_input, translate_operation->get_input_socket(0));
converter.add_link(translate_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.add_link(rotate_operation->get_output_socket(),
scale_operation->get_input_socket(0));
converter.map_output_socket(get_output_socket(), scale_operation->get_output_socket());
}
else {
/* Scale -> Rotate -> Translate. */
stabilization_socket = scale_operation->get_input_socket(0);
converter.map_input_socket(image_input, scale_operation->get_input_socket(0));
converter.add_link(scale_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.add_link(rotate_operation->get_output_socket(),
translate_operation->get_input_socket(0));
converter.map_output_socket(get_output_socket(), translate_operation->get_output_socket());
}
x_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
y_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
scale_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
angle_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
break;
}
converter.map_output_socket(get_output_socket(), scale_operation->get_output_socket());
}
else {
/* Scale -> Rotate -> Translate. */
stabilization_socket = scale_operation->get_input_socket(0);
converter.map_input_socket(image_input, scale_operation->get_input_socket(0));
converter.add_link(scale_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.add_link(rotate_operation->get_output_socket(),
translate_operation->get_input_socket(0));
converter.map_output_socket(get_output_socket(), translate_operation->get_output_socket());
}
x_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
y_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
scale_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
angle_attribute->set_socket_input_resolution_for_stabilization(stabilization_socket);
}
} // namespace blender::compositor

83
source/blender/compositor/nodes/COM_TransformNode.cc
View File

@ -16,7 +16,7 @@ TransformNode::TransformNode(bNode *editor_node) : Node(editor_node)
}
void TransformNode::convert_to_operations(NodeConverter &converter,
const CompositorContext &context) const
const CompositorContext & /*context*/) const
{
NodeInput *image_input = this->get_input_socket(0);
NodeInput *x_input = this->get_input_socket(1);
@ -24,73 +24,34 @@ void TransformNode::convert_to_operations(NodeConverter &converter,
NodeInput *angle_input = this->get_input_socket(3);
NodeInput *scale_input = this->get_input_socket(4);
switch (context.get_execution_model()) {
case eExecutionModel::Tiled: {
ScaleRelativeOperation *scale_operation = new ScaleRelativeOperation();
converter.add_operation(scale_operation);
ScaleRelativeOperation *scale_operation = new ScaleRelativeOperation();
converter.add_operation(scale_operation);
RotateOperation *rotate_operation = new RotateOperation();
rotate_operation->set_do_degree2_rad_conversion(false);
converter.add_operation(rotate_operation);
RotateOperation *rotate_operation = new RotateOperation();
rotate_operation->set_do_degree2_rad_conversion(false);
converter.add_operation(rotate_operation);
TranslateOperation *translate_operation = new TranslateOperation();
converter.add_operation(translate_operation);
TranslateOperation *translate_operation = new TranslateCanvasOperation();
converter.add_operation(translate_operation);
SetSamplerOperation *sampler = new SetSamplerOperation();
sampler->set_sampler((PixelSampler)this->get_bnode()->custom1);
converter.add_operation(sampler);
PixelSampler sampler = (PixelSampler)this->get_bnode()->custom1;
scale_operation->set_sampler(sampler);
rotate_operation->set_sampler(sampler);
converter.map_input_socket(image_input, sampler->get_input_socket(0));
converter.add_link(sampler->get_output_socket(), scale_operation->get_input_socket(0));
converter.map_input_socket(scale_input, scale_operation->get_input_socket(1));
converter.map_input_socket(scale_input,
scale_operation->get_input_socket(2)); // xscale = yscale
converter.map_input_socket(image_input, scale_operation->get_input_socket(0));
converter.map_input_socket(scale_input, scale_operation->get_input_socket(1));
converter.map_input_socket(scale_input,
scale_operation->get_input_socket(2)); // xscale = yscale
converter.add_link(scale_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.map_input_socket(angle_input, rotate_operation->get_input_socket(1));
converter.add_link(scale_operation->get_output_socket(), rotate_operation->get_input_socket(0));
converter.map_input_socket(angle_input, rotate_operation->get_input_socket(1));
converter.add_link(rotate_operation->get_output_socket(),
translate_operation->get_input_socket(0));
converter.map_input_socket(x_input, translate_operation->get_input_socket(1));
converter.map_input_socket(y_input, translate_operation->get_input_socket(2));
converter.add_link(rotate_operation->get_output_socket(),
translate_operation->get_input_socket(0));
converter.map_input_socket(x_input, translate_operation->get_input_socket(1));
converter.map_input_socket(y_input, translate_operation->get_input_socket(2));
converter.map_output_socket(get_output_socket(), translate_operation->get_output_socket());
break;
}
case eExecutionModel::FullFrame: {
ScaleRelativeOperation *scale_operation = new ScaleRelativeOperation();
converter.add_operation(scale_operation);
RotateOperation *rotate_operation = new RotateOperation();
rotate_operation->set_do_degree2_rad_conversion(false);
converter.add_operation(rotate_operation);
TranslateOperation *translate_operation = new TranslateCanvasOperation();
converter.add_operation(translate_operation);
PixelSampler sampler = (PixelSampler)this->get_bnode()->custom1;
scale_operation->set_sampler(sampler);
rotate_operation->set_sampler(sampler);
converter.map_input_socket(image_input, scale_operation->get_input_socket(0));
converter.map_input_socket(scale_input, scale_operation->get_input_socket(1));
converter.map_input_socket(scale_input,
scale_operation->get_input_socket(2)); // xscale = yscale
converter.add_link(scale_operation->get_output_socket(),
rotate_operation->get_input_socket(0));
converter.map_input_socket(angle_input, rotate_operation->get_input_socket(1));
converter.add_link(rotate_operation->get_output_socket(),
translate_operation->get_input_socket(0));
converter.map_input_socket(x_input, translate_operation->get_input_socket(1));
converter.map_input_socket(y_input, translate_operation->get_input_socket(2));
converter.map_output_socket(get_output_socket(), translate_operation->get_output_socket());
break;
}
}
converter.map_output_socket(get_output_socket(), translate_operation->get_output_socket());
}
} // namespace blender::compositor

24
source/blender/compositor/nodes/COM_TranslateNode.cc
View File

@ -5,8 +5,6 @@
#include "COM_TranslateNode.h"
#include "COM_TranslateOperation.h"
#include "COM_WrapOperation.h"
#include "COM_WriteBufferOperation.h"
namespace blender::compositor {
@ -16,7 +14,7 @@ TranslateNode::TranslateNode(bNode *editor_node) : Node(editor_node)
}
void TranslateNode::convert_to_operations(NodeConverter &converter,
const CompositorContext &context) const
const CompositorContext & /*context*/) const
{
const bNode *bnode = this->get_bnode();
const NodeTranslateData *data = (const NodeTranslateData *)bnode->storage;
@ -26,9 +24,7 @@ void TranslateNode::convert_to_operations(NodeConverter &converter,
NodeInput *input_ysocket = this->get_input_socket(2);
NodeOutput *output_socket = this->get_output_socket(0);
TranslateOperation *operation = context.get_execution_model() == eExecutionModel::Tiled ?
new TranslateOperation() :
new TranslateCanvasOperation();
TranslateOperation *operation = new TranslateCanvasOperation();
operation->set_wrapping(data->wrap_axis);
operation->set_is_relative(data->relative);
@ -36,21 +32,7 @@ void TranslateNode::convert_to_operations(NodeConverter &converter,
converter.map_input_socket(input_xsocket, operation->get_input_socket(1));
converter.map_input_socket(input_ysocket, operation->get_input_socket(2));
converter.map_output_socket(output_socket, operation->get_output_socket(0));
if (data->wrap_axis && context.get_execution_model() != eExecutionModel::FullFrame) {
/* TODO: To be removed with tiled implementation. */
WriteBufferOperation *write_operation = new WriteBufferOperation(DataType::Color);
WrapOperation *wrap_operation = new WrapOperation(DataType::Color);
wrap_operation->set_memory_proxy(write_operation->get_memory_proxy());
wrap_operation->set_wrapping(data->wrap_axis);
converter.add_operation(write_operation);
converter.add_operation(wrap_operation);
converter.map_input_socket(input_socket, write_operation->get_input_socket(0));
converter.add_link(wrap_operation->get_output_socket(), operation->get_input_socket(0));
}
else {
converter.map_input_socket(input_socket, operation->get_input_socket(0));
}
converter.map_input_socket(input_socket, operation->get_input_socket(0));
}
} // namespace blender::compositor

1
source/blender/compositor/nodes/COM_ViewerNode.cc
View File

@ -29,7 +29,6 @@ void ViewerNode::convert_to_operations(NodeConverter &converter,
viewer_operation->set_bnodetree(context.get_bnodetree());
viewer_operation->set_image(image);
viewer_operation->set_image_user(image_user);
viewer_operation->set_chunk_order((ChunkOrdering)editor_node->custom1);
viewer_operation->setCenterX(editor_node->custom3);
viewer_operation->setCenterY(editor_node->custom4);
/* alpha socket gives either 1 or a custom alpha value if "use alpha" is enabled */

30
source/blender/compositor/operations/COM_AlphaOverKeyOperation.cc
View File

@ -11,36 +11,6 @@ AlphaOverKeyOperation::AlphaOverKeyOperation()
flags_.can_be_constant = true;
}
void AlphaOverKeyOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_color1[4];
float input_over_color[4];
float value[4];
input_value_operation_->read_sampled(value, x, y, sampler);
input_color1_operation_->read_sampled(input_color1, x, y, sampler);
input_color2_operation_->read_sampled(input_over_color, x, y, sampler);
if (input_over_color[3] <= 0.0f) {
copy_v4_v4(output, input_color1);
}
else if (value[0] == 1.0f && input_over_color[3] >= 1.0f) {
copy_v4_v4(output, input_over_color);
}
else {
float premul = value[0] * input_over_color[3];
float mul = 1.0f - premul;
output[0] = (mul * input_color1[0]) + premul * input_over_color[0];
output[1] = (mul * input_color1[1]) + premul * input_over_color[1];
output[2] = (mul * input_color1[2]) + premul * input_over_color[2];
output[3] = (mul * input_color1[3]) + value[0] * input_over_color[3];
}
}
void AlphaOverKeyOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {

5
source/blender/compositor/operations/COM_AlphaOverKeyOperation.h
View File

@ -16,11 +16,6 @@ class AlphaOverKeyOperation : public MixBaseOperation {
public:
AlphaOverKeyOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void update_memory_buffer_row(PixelCursor &p) override;
};

31
source/blender/compositor/operations/COM_AlphaOverMixedOperation.cc
View File

@ -12,37 +12,6 @@ AlphaOverMixedOperation::AlphaOverMixedOperation()
flags_.can_be_constant = true;
}
void AlphaOverMixedOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_color1[4];
float input_over_color[4];
float value[4];
input_value_operation_->read_sampled(value, x, y, sampler);
input_color1_operation_->read_sampled(input_color1, x, y, sampler);
input_color2_operation_->read_sampled(input_over_color, x, y, sampler);
if (input_over_color[3] <= 0.0f) {
copy_v4_v4(output, input_color1);
}
else if (value[0] == 1.0f && input_over_color[3] >= 1.0f) {
copy_v4_v4(output, input_over_color);
}
else {
float addfac = 1.0f - x_ + input_over_color[3] * x_;
float premul = value[0] * addfac;
float mul = 1.0f - value[0] * input_over_color[3];
output[0] = (mul * input_color1[0]) + premul * input_over_color[0];
output[1] = (mul * input_color1[1]) + premul * input_over_color[1];
output[2] = (mul * input_color1[2]) + premul * input_over_color[2];
output[3] = (mul * input_color1[3]) + value[0] * input_over_color[3];
}
}
void AlphaOverMixedOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {

5
source/blender/compositor/operations/COM_AlphaOverMixedOperation.h
View File

@ -22,11 +22,6 @@ class AlphaOverMixedOperation : public MixBaseOperation {
*/
AlphaOverMixedOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void setX(float x)
{
x_ = x;

30
source/blender/compositor/operations/COM_AlphaOverPremultiplyOperation.cc
View File

@ -11,36 +11,6 @@ AlphaOverPremultiplyOperation::AlphaOverPremultiplyOperation()
flags_.can_be_constant = true;
}
void AlphaOverPremultiplyOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_color1[4];
float input_over_color[4];
float value[4];
input_value_operation_->read_sampled(value, x, y, sampler);
input_color1_operation_->read_sampled(input_color1, x, y, sampler);
input_color2_operation_->read_sampled(input_over_color, x, y, sampler);
/* Zero alpha values should still permit an add of RGB data */
if (input_over_color[3] < 0.0f) {
copy_v4_v4(output, input_color1);
}
else if (value[0] == 1.0f && input_over_color[3] >= 1.0f) {
copy_v4_v4(output, input_over_color);
}
else {
float mul = 1.0f - value[0] * input_over_color[3];
output[0] = (mul * input_color1[0]) + value[0] * input_over_color[0];
output[1] = (mul * input_color1[1]) + value[0] * input_over_color[1];
output[2] = (mul * input_color1[2]) + value[0] * input_over_color[2];
output[3] = (mul * input_color1[3]) + value[0] * input_over_color[3];
}
}
void AlphaOverPremultiplyOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {

5
source/blender/compositor/operations/COM_AlphaOverPremultiplyOperation.h
View File

@ -16,11 +16,6 @@ class AlphaOverPremultiplyOperation : public MixBaseOperation {
public:
AlphaOverPremultiplyOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void update_memory_buffer_row(PixelCursor &p) override;
};

73
source/blender/compositor/operations/COM_BilateralBlurOperation.cc
View File

@ -11,87 +11,14 @@ BilateralBlurOperation::BilateralBlurOperation()
this->add_input_socket(DataType::Color);
this->add_input_socket(DataType::Color);
this->add_output_socket(DataType::Color);
flags_.complex = true;
flags_.can_be_constant = true;
input_color_program_ = nullptr;
input_determinator_program_ = nullptr;
}
void BilateralBlurOperation::init_execution()
{
input_color_program_ = get_input_socket_reader(0);
input_determinator_program_ = get_input_socket_reader(1);
QualityStepHelper::init_execution(COM_QH_INCREASE);
}
void BilateralBlurOperation::execute_pixel(float output[4], int x, int y, void *data)
{
/* Read the determinator color at x, y,
* this will be used as the reference color for the determinator. */
float determinator_reference_color[4];
float determinator[4];
float temp_color[4];
float blur_color[4];
float blur_divider;
float sigmacolor = data_->sigma_color;
float delta_color;
input_determinator_program_->read(determinator_reference_color, x, y, data);
zero_v4(blur_color);
blur_divider = 0.0f;
/* TODO(sergey): This isn't really good bilateral filter, it should be
* using gaussian bell for weights. Also sigma_color doesn't seem to be
* used correct at all.
*/
for (int yi = -radius_; yi <= radius_; yi += QualityStepHelper::get_step()) {
for (int xi = -radius_; xi <= radius_; xi += QualityStepHelper::get_step()) {
/* Read determinator. */
input_determinator_program_->read_clamped(determinator, x + xi, y + yi, data);
delta_color = (fabsf(determinator_reference_color[0] - determinator[0]) +
fabsf(determinator_reference_color[1] - determinator[1]) +
/* Do not take the alpha channel into account. */
fabsf(determinator_reference_color[2] - determinator[2]));
if (delta_color < sigmacolor) {
/* Add this to the blur. */
input_color_program_->read_clamped(temp_color, x + xi, y + yi, data);
add_v4_v4(blur_color, temp_color);
blur_divider += 1.0f;
}
}
}
if (blur_divider > 0.0f) {
mul_v4_v4fl(output, blur_color, 1.0f / blur_divider);
}
else {
output[0] = 0.0f;
output[1] = 0.0f;
output[2] = 0.0f;
output[3] = 1.0f;
}
}
void BilateralBlurOperation::deinit_execution()
{
input_color_program_ = nullptr;
input_determinator_program_ = nullptr;
}
bool BilateralBlurOperation::determine_depending_area_of_interest(
rcti *input, ReadBufferOperation *read_operation, rcti *output)
{
rcti new_input;
int add = radius_ + 1;
new_input.xmax = input->xmax + (add);
new_input.xmin = input->xmin - (add);
new_input.ymax = input->ymax + (add);
new_input.ymin = input->ymin - (add);
return NodeOperation::determine_depending_area_of_interest(&new_input, read_operation, output);
}
void BilateralBlurOperation::get_area_of_interest(const int /*input_idx*/,
const rcti &output_area,
rcti &r_input_area)

19
source/blender/compositor/operations/COM_BilateralBlurOperation.h
View File

@ -13,33 +13,14 @@ namespace blender::compositor {
class BilateralBlurOperation : public MultiThreadedOperation, public QualityStepHelper {
private:
SocketReader *input_color_program_;
SocketReader *input_determinator_program_;
NodeBilateralBlurData *data_;
int radius_;
public:
BilateralBlurOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel(float output[4], int x, int y, void *data) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
bool determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output) override;
void set_data(NodeBilateralBlurData *data)
{
data_ = data;

63
source/blender/compositor/operations/COM_BlurBaseOperation.cc
View File

@ -15,9 +15,7 @@ BlurBaseOperation::BlurBaseOperation(DataType data_type)
this->add_input_socket(data_type);
this->add_input_socket(DataType::Value);
this->add_output_socket(data_type);
flags_.complex = true;
flags_.can_be_constant = true;
input_program_ = nullptr;
memset(&data_, 0, sizeof(NodeBlurData));
size_ = 1.0f;
sizeavailable_ = false;
@ -27,9 +25,7 @@ BlurBaseOperation::BlurBaseOperation(DataType data_type)
void BlurBaseOperation::init_data()
{
if (execution_model_ == eExecutionModel::FullFrame) {
update_size();
}
update_size();
data_.image_in_width = this->get_width();
data_.image_in_height = this->get_height();
@ -55,9 +51,6 @@ void BlurBaseOperation::init_data()
void BlurBaseOperation::init_execution()
{
input_program_ = this->get_input_socket_reader(0);
input_size_ = this->get_input_socket_reader(1);
QualityStepHelper::init_execution(COM_QH_MULTIPLY);
}
@ -148,12 +141,6 @@ float *BlurBaseOperation::make_dist_fac_inverse(float rad, int size, int falloff
return dist_fac_invert;
}
void BlurBaseOperation::deinit_execution()
{
input_program_ = nullptr;
input_size_ = nullptr;
}
void BlurBaseOperation::set_data(const NodeBlurData *data)
{
memcpy(&data_, data, sizeof(NodeBlurData));
@ -176,21 +163,10 @@ void BlurBaseOperation::update_size()
return;
}
switch (execution_model_) {
case eExecutionModel::Tiled: {
float result[4];
this->get_input_socket_reader(1)->read_sampled(result, 0, 0, PixelSampler::Nearest);
size_ = result[0];
break;
}
case eExecutionModel::FullFrame: {
NodeOperation *size_input = get_input_operation(SIZE_INPUT_INDEX);
if (size_input->get_flags().is_constant_operation) {
size_ = *static_cast<ConstantOperation *>(size_input)->get_constant_elem();
} /* Else use default. */
break;
}
}
NodeOperation *size_input = get_input_operation(SIZE_INPUT_INDEX);
if (size_input->get_flags().is_constant_operation) {
size_ = *static_cast<ConstantOperation *>(size_input)->get_constant_elem();
} /* Else use default. */
sizeavailable_ = true;
}
@ -201,26 +177,15 @@ void BlurBaseOperation::determine_canvas(const rcti &preferred_area, rcti &r_are
return;
}
switch (execution_model_) {
case eExecutionModel::Tiled: {
NodeOperation::determine_canvas(preferred_area, r_area);
r_area.xmax += 2 * size_ * data_.sizex;
r_area.ymax += 2 * size_ * data_.sizey;
break;
}
case eExecutionModel::FullFrame: {
/* Setting a modifier ensures all non main inputs have extended bounds as preferred
* canvas, avoiding unnecessary canvas conversions that would hide constant
* operations. */
set_determined_canvas_modifier([=](rcti &canvas) {
/* Rounding to even prevents jiggling in backdrop while switching size values. */
canvas.xmax += round_to_even(2 * size_ * data_.sizex);
canvas.ymax += round_to_even(2 * size_ * data_.sizey);
});
NodeOperation::determine_canvas(preferred_area, r_area);
break;
}
}
/* Setting a modifier ensures all non main inputs have extended bounds as preferred
* canvas, avoiding unnecessary canvas conversions that would hide constant
* operations. */
set_determined_canvas_modifier([=](rcti &canvas) {
/* Rounding to even prevents jiggling in backdrop while switching size values. */
canvas.xmax += round_to_even(2 * size_ * data_.sizex);
canvas.ymax += round_to_even(2 * size_ * data_.sizey);
});
NodeOperation::determine_canvas(preferred_area, r_area);
}
void BlurBaseOperation::get_area_of_interest(const int input_idx,

13
source/blender/compositor/operations/COM_BlurBaseOperation.h
View File

@ -35,11 +35,6 @@ class BlurBaseOperation : public MultiThreadedOperation, public QualityStepHelpe
void update_size();
/**
* Cached reference to the input_program
*/
SocketReader *input_program_;
SocketReader *input_size_;
NodeBlurData data_;
float size_;
@ -50,16 +45,8 @@ class BlurBaseOperation : public MultiThreadedOperation, public QualityStepHelpe
public:
virtual void init_data() override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_data(const NodeBlurData *data);
void set_size(float size)

203
source/blender/compositor/operations/COM_BokehBlurOperation.cc
View File

@ -8,8 +8,6 @@
#include "COM_BokehBlurOperation.h"
#include "COM_ConstantOperation.h"
#include "COM_OpenCLDevice.h"
namespace blender::compositor {
constexpr int IMAGE_INPUT_INDEX = 0;
@ -25,195 +23,35 @@ BokehBlurOperation::BokehBlurOperation()
this->add_input_socket(DataType::Value);
this->add_output_socket(DataType::Color);
flags_.complex = true;
flags_.open_cl = true;
flags_.can_be_constant = true;
size_ = 1.0f;
sizeavailable_ = false;
input_program_ = nullptr;
input_bokeh_program_ = nullptr;
input_bounding_box_reader_ = nullptr;
extend_bounds_ = false;
}
void BokehBlurOperation::init_data()
{
if (execution_model_ == eExecutionModel::FullFrame) {
update_size();
}
}
void *BokehBlurOperation::initialize_tile_data(rcti * /*rect*/)
{
lock_mutex();
if (!sizeavailable_) {
update_size();
}
void *buffer = get_input_operation(0)->initialize_tile_data(nullptr);
unlock_mutex();
return buffer;
update_size();
}
void BokehBlurOperation::init_execution()
{
init_mutex();
input_program_ = get_input_socket_reader(0);
input_bokeh_program_ = get_input_socket_reader(1);
input_bounding_box_reader_ = get_input_socket_reader(2);
QualityStepHelper::init_execution(COM_QH_INCREASE);
}
void BokehBlurOperation::execute_pixel(float output[4], int x, int y, void *data)
{
MemoryBuffer *input_buffer = (MemoryBuffer *)data;
float temp_bounding_box[4];
input_bounding_box_reader_->read_sampled(temp_bounding_box, x, y, PixelSampler::Nearest);
if (temp_bounding_box[0] <= 0.0f) {
copy_v4_v4(output, input_buffer->get_elem(x, y));
return;
}
const float max_dim = std::max(this->get_width(), this->get_height());
int radius = size_ * max_dim / 100.0f;
const int2 bokeh_size = int2(input_bokeh_program_->get_width(),
input_bokeh_program_->get_height());
float4 accumulated_color = float4(0.0f);
float4 accumulated_weight = float4(0.0f);
int step = get_step();
for (int yi = -radius; yi <= radius; yi += step) {
for (int xi = -radius; xi <= radius; xi += step) {
const float2 normalized_texel = (float2(xi, yi) + radius + 0.5f) / (radius * 2.0f + 1.0f);
const float2 weight_texel = (1.0f - normalized_texel) * float2(bokeh_size - 1);
float4 weight;
input_bokeh_program_->read(weight, int(weight_texel.x), int(weight_texel.y), nullptr);
const float4 color = float4(input_buffer->get_elem_clamped(x + xi, y + yi)) * weight;
accumulated_color += color;
accumulated_weight += weight;
}
}
const float4 final_color = math::safe_divide(accumulated_color, accumulated_weight);
copy_v4_v4(output, final_color);
}
void BokehBlurOperation::deinit_execution()
{
deinit_mutex();
input_program_ = nullptr;
input_bokeh_program_ = nullptr;
input_bounding_box_reader_ = nullptr;
}
bool BokehBlurOperation::determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output)
{
rcti new_input;
rcti bokeh_input;
const float max_dim = std::max(this->get_width(), this->get_height());
if (sizeavailable_) {
new_input.xmax = input->xmax + (size_ * max_dim / 100.0f);
new_input.xmin = input->xmin - (size_ * max_dim / 100.0f);
new_input.ymax = input->ymax + (size_ * max_dim / 100.0f);
new_input.ymin = input->ymin - (size_ * max_dim / 100.0f);
}
else {
new_input.xmax = input->xmax + (10.0f * max_dim / 100.0f);
new_input.xmin = input->xmin - (10.0f * max_dim / 100.0f);
new_input.ymax = input->ymax + (10.0f * max_dim / 100.0f);
new_input.ymin = input->ymin - (10.0f * max_dim / 100.0f);
}
NodeOperation *operation = get_input_operation(1);
bokeh_input.xmax = operation->get_width();
bokeh_input.xmin = 0;
bokeh_input.ymax = operation->get_height();
bokeh_input.ymin = 0;
if (operation->determine_depending_area_of_interest(&bokeh_input, read_operation, output)) {
return true;
}
operation = get_input_operation(0);
if (operation->determine_depending_area_of_interest(&new_input, read_operation, output)) {
return true;
}
operation = get_input_operation(2);
if (operation->determine_depending_area_of_interest(input, read_operation, output)) {
return true;
}
if (!sizeavailable_) {
rcti size_input;
size_input.xmin = 0;
size_input.ymin = 0;
size_input.xmax = 5;
size_input.ymax = 5;
operation = get_input_operation(3);
if (operation->determine_depending_area_of_interest(&size_input, read_operation, output)) {
return true;
}
}
return false;
}
void BokehBlurOperation::execute_opencl(OpenCLDevice *device,
MemoryBuffer *output_memory_buffer,
cl_mem cl_output_buffer,
MemoryBuffer **input_memory_buffers,
std::list<cl_mem> *cl_mem_to_clean_up,
std::list<cl_kernel> * /*cl_kernels_to_clean_up*/)
{
cl_kernel kernel = device->COM_cl_create_kernel("bokeh_blur_kernel", nullptr);
if (!sizeavailable_) {
update_size();
}
const float max_dim = std::max(this->get_width(), this->get_height());
cl_int radius = size_ * max_dim / 100.0f;
cl_int step = this->get_step();
device->COM_cl_attach_memory_buffer_to_kernel_parameter(
kernel, 0, -1, cl_mem_to_clean_up, input_memory_buffers, input_bounding_box_reader_);
device->COM_cl_attach_memory_buffer_to_kernel_parameter(
kernel, 1, 4, cl_mem_to_clean_up, input_memory_buffers, input_program_);
device->COM_cl_attach_memory_buffer_to_kernel_parameter(
kernel, 2, -1, cl_mem_to_clean_up, input_memory_buffers, input_bokeh_program_);
device->COM_cl_attach_output_memory_buffer_to_kernel_parameter(kernel, 3, cl_output_buffer);
device->COM_cl_attach_memory_buffer_offset_to_kernel_parameter(kernel, 5, output_memory_buffer);
clSetKernelArg(kernel, 6, sizeof(cl_int), &radius);
clSetKernelArg(kernel, 7, sizeof(cl_int), &step);
device->COM_cl_attach_size_to_kernel_parameter(kernel, 8, this);
device->COM_cl_enqueue_range(kernel, output_memory_buffer, 9, this);
}
void BokehBlurOperation::update_size()
{
if (sizeavailable_) {
return;
}
switch (execution_model_) {
case eExecutionModel::Tiled: {
float result[4];
this->get_input_socket_reader(3)->read_sampled(result, 0, 0, PixelSampler::Nearest);
size_ = result[0];
CLAMP(size_, 0.0f, 10.0f);
break;
}
case eExecutionModel::FullFrame: {
NodeOperation *size_input = get_input_operation(SIZE_INPUT_INDEX);
if (size_input->get_flags().is_constant_operation) {
size_ = *static_cast<ConstantOperation *>(size_input)->get_constant_elem();
CLAMP(size_, 0.0f, 10.0f);
} /* Else use default. */
break;
}
}
NodeOperation *size_input = get_input_operation(SIZE_INPUT_INDEX);
if (size_input->get_flags().is_constant_operation) {
size_ = *static_cast<ConstantOperation *>(size_input)->get_constant_elem();
CLAMP(size_, 0.0f, 10.0f);
} /* Else use default. */
sizeavailable_ = true;
}
@ -224,27 +62,14 @@ void BokehBlurOperation::determine_canvas(const rcti &preferred_area, rcti &r_ar
return;
}
switch (execution_model_) {
case eExecutionModel::Tiled: {
NodeOperation::determine_canvas(preferred_area, r_area);
const float max_dim = std::max(BLI_rcti_size_x(&r_area), BLI_rcti_size_y(&r_area));
float add_size = round_to_even(2 * size_ * max_dim / 100.0f);
r_area.xmax += add_size;
r_area.ymax += add_size;
break;
}
case eExecutionModel::FullFrame: {
set_determined_canvas_modifier([=](rcti &canvas) {
const float max_dim = std::max(BLI_rcti_size_x(&canvas), BLI_rcti_size_y(&canvas));
/* Rounding to even prevents image jiggling in backdrop while switching size values. */
float add_size = round_to_even(2 * size_ * max_dim / 100.0f);
canvas.xmax += add_size;
canvas.ymax += add_size;
});
NodeOperation::determine_canvas(preferred_area, r_area);
break;
}
}
set_determined_canvas_modifier([=](rcti &canvas) {
const float max_dim = std::max(BLI_rcti_size_x(&canvas), BLI_rcti_size_y(&canvas));
/* Rounding to even prevents image jiggling in backdrop while switching size values. */
float add_size = round_to_even(2 * size_ * max_dim / 100.0f);
canvas.xmax += add_size;
canvas.ymax += add_size;
});
NodeOperation::determine_canvas(preferred_area, r_area);
}
void BokehBlurOperation::get_area_of_interest(const int input_idx,

28
source/blender/compositor/operations/COM_BokehBlurOperation.h
View File

@ -11,9 +11,6 @@ namespace blender::compositor {
class BokehBlurOperation : public MultiThreadedOperation, public QualityStepHelper {
private:
SocketReader *input_program_;
SocketReader *input_bokeh_program_;
SocketReader *input_bounding_box_reader_;
void update_size();
float size_;
bool sizeavailable_;
@ -25,39 +22,14 @@ class BokehBlurOperation : public MultiThreadedOperation, public QualityStepHelp
void init_data() override;
void *initialize_tile_data(rcti *rect) override;
/**
* The inner loop of this operation.
*/
void execute_pixel(float output[4], int x, int y, void *data) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
bool determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output) override;
void set_size(float size)
{
size_ = size;
sizeavailable_ = true;
}
void execute_opencl(OpenCLDevice *device,
MemoryBuffer *output_memory_buffer,
cl_mem cl_output_buffer,
MemoryBuffer **input_memory_buffers,
std::list<cl_mem> *cl_mem_to_clean_up,
std::list<cl_kernel> *cl_kernels_to_clean_up) override;
void set_extend_bounds(bool extend_bounds)
{
extend_bounds_ = extend_bounds;

24
source/blender/compositor/operations/COM_BokehImageOperation.cc
View File

@ -75,30 +75,6 @@ float BokehImageOperation::is_inside_bokeh(float distance, float x, float y)
}
return inside_bokeh;
}
void BokehImageOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler /*sampler*/)
{
float shift = data_->lensshift;
float shift2 = shift / 2.0f;
float distance = circular_distance_;
float inside_bokeh_max = is_inside_bokeh(distance, x, y);
float inside_bokeh_med = is_inside_bokeh(distance - fabsf(shift2 * distance), x, y);
float inside_bokeh_min = is_inside_bokeh(distance - fabsf(shift * distance), x, y);
if (shift < 0) {
output[0] = inside_bokeh_max;
output[1] = inside_bokeh_med;
output[2] = inside_bokeh_min;
}
else {
output[0] = inside_bokeh_min;
output[1] = inside_bokeh_med;
output[2] = inside_bokeh_max;
}
output[3] = (inside_bokeh_max + inside_bokeh_med + inside_bokeh_min) / 3.0f;
}
void BokehImageOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> /*inputs*/)

12
source/blender/compositor/operations/COM_BokehImageOperation.h
View File

@ -95,19 +95,7 @@ class BokehImageOperation : public MultiThreadedOperation {
public:
BokehImageOperation();
/**
* \brief The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* \brief Initialize the execution
*/
void init_execution() override;
/**
* \brief De-initialize the execution
*/
void deinit_execution() override;
/**

76
source/blender/compositor/operations/COM_BoxMaskOperation.cc
View File

@ -11,87 +11,17 @@ BoxMaskOperation::BoxMaskOperation()
this->add_input_socket(DataType::Value);
this->add_input_socket(DataType::Value);
this->add_output_socket(DataType::Value);
input_mask_ = nullptr;
input_value_ = nullptr;
cosine_ = 0.0f;
sine_ = 0.0f;
}
void BoxMaskOperation::init_execution()
{
input_mask_ = this->get_input_socket_reader(0);
input_value_ = this->get_input_socket_reader(1);
const double rad = double(data_->rotation);
cosine_ = cos(rad);
sine_ = sin(rad);
aspect_ratio_ = float(this->get_width()) / this->get_height();
}
void BoxMaskOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_mask[4];
float input_value[4];
float rx = x / std::max(this->get_width() - 1.0f, FLT_EPSILON);
float ry = y / std::max(this->get_height() - 1.0f, FLT_EPSILON);
const float dy = (ry - data_->y) / aspect_ratio_;
const float dx = rx - data_->x;
rx = data_->x + (cosine_ * dx + sine_ * dy);
ry = data_->y + (-sine_ * dx + cosine_ * dy);
input_mask_->read_sampled(input_mask, x, y, sampler);
input_value_->read_sampled(input_value, x, y, sampler);
float half_height = data_->height / 2.0f + FLT_EPSILON;
float half_width = data_->width / 2.0f + FLT_EPSILON;
bool inside = (rx >= data_->x - half_width && rx <= data_->x + half_width &&
ry >= data_->y - half_height && ry <= data_->y + half_height);
switch (mask_type_) {
case CMP_NODE_MASKTYPE_ADD:
if (inside) {
output[0] = std::max(input_mask[0], input_value[0]);
}
else {
output[0] = input_mask[0];
}
break;
case CMP_NODE_MASKTYPE_SUBTRACT:
if (inside) {
output[0] = input_mask[0] - input_value[0];
CLAMP(output[0], 0, 1);
}
else {
output[0] = input_mask[0];
}
break;
case CMP_NODE_MASKTYPE_MULTIPLY:
if (inside) {
output[0] = input_mask[0] * input_value[0];
}
else {
output[0] = 0;
}
break;
case CMP_NODE_MASKTYPE_NOT:
if (inside) {
if (input_mask[0] > 0.0f) {
output[0] = 0;
}
else {
output[0] = input_value[0];
}
}
else {
output[0] = input_mask[0];
}
break;
}
}
void BoxMaskOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)
@ -150,10 +80,4 @@ void BoxMaskOperation::apply_mask(MemoryBuffer *output,
}
}
void BoxMaskOperation::deinit_execution()
{
input_mask_ = nullptr;
input_value_ = nullptr;
}
} // namespace blender::compositor

19
source/blender/compositor/operations/COM_BoxMaskOperation.h
View File

@ -12,12 +12,6 @@ class BoxMaskOperation : public MultiThreadedOperation {
private:
using MaskFunc = std::function<float(bool is_inside, const float *mask, const float *value)>;
/**
* Cached reference to the input_program
*/
SocketReader *input_mask_;
SocketReader *input_value_;
float sine_;
float cosine_;
float aspect_ratio_;
@ -28,21 +22,8 @@ class BoxMaskOperation : public MultiThreadedOperation {
public:
BoxMaskOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_data(const NodeBoxMask *data)
{
data_ = data;

58
source/blender/compositor/operations/COM_BrightnessOperation.cc
View File

@ -14,7 +14,6 @@ BrightnessOperation::BrightnessOperation()
this->add_input_socket(DataType::Value);
this->add_input_socket(DataType::Value);
this->add_output_socket(DataType::Color);
input_program_ = nullptr;
use_premultiply_ = false;
flags_.can_be_constant = true;
}
@ -24,56 +23,6 @@ void BrightnessOperation::set_use_premultiply(bool use_premultiply)
use_premultiply_ = use_premultiply;
}
void BrightnessOperation::init_execution()
{
input_program_ = this->get_input_socket_reader(0);
input_brightness_program_ = this->get_input_socket_reader(1);
input_contrast_program_ = this->get_input_socket_reader(2);
}
void BrightnessOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_value[4];
float a, b;
float input_brightness[4];
float input_contrast[4];
input_program_->read_sampled(input_value, x, y, sampler);
input_brightness_program_->read_sampled(input_brightness, x, y, sampler);
input_contrast_program_->read_sampled(input_contrast, x, y, sampler);
float brightness = input_brightness[0];
float contrast = input_contrast[0];
brightness /= 100.0f;
float delta = contrast / 200.0f;
/*
* The algorithm is by Werner D. Streidt
* (http://visca.com/ffactory/archives/5-99/msg00021.html)
* Extracted of OpenCV `demhist.c`.
*/
if (contrast > 0) {
a = 1.0f - delta * 2.0f;
a = 1.0f / max_ff(a, FLT_EPSILON);
b = a * (brightness - delta);
}
else {
delta *= -1;
a = max_ff(1.0f - delta * 2.0f, 0.0f);
b = a * brightness + delta;
}
if (use_premultiply_) {
premul_to_straight_v4(input_value);
}
output[0] = a * input_value[0] + b;
output[1] = a * input_value[1] + b;
output[2] = a * input_value[2] + b;
output[3] = input_value[3];
if (use_premultiply_) {
straight_to_premul_v4(output);
}
}
void BrightnessOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)
@ -118,11 +67,4 @@ void BrightnessOperation::update_memory_buffer_partial(MemoryBuffer *output,
}
}
void BrightnessOperation::deinit_execution()
{
input_program_ = nullptr;
input_brightness_program_ = nullptr;
input_contrast_program_ = nullptr;
}
} // namespace blender::compositor

22
source/blender/compositor/operations/COM_BrightnessOperation.h
View File

@ -10,33 +10,11 @@ namespace blender::compositor {
class BrightnessOperation : public MultiThreadedOperation {
private:
/**
* Cached reference to the input_program
*/
SocketReader *input_program_;
SocketReader *input_brightness_program_;
SocketReader *input_contrast_program_;
bool use_premultiply_;
public:
BrightnessOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_use_premultiply(bool use_premultiply);
void update_memory_buffer_partial(MemoryBuffer *output,

90
source/blender/compositor/operations/COM_CalculateMeanOperation.cc
View File

@ -14,10 +14,8 @@ CalculateMeanOperation::CalculateMeanOperation()
{
this->add_input_socket(DataType::Color, ResizeMode::Align);
this->add_output_socket(DataType::Value);
image_reader_ = nullptr;
is_calculated_ = false;
setting_ = 1;
flags_.complex = true;
flags_.is_constant_operation = true;
needs_canvas_to_get_constant_ = true;
@ -25,95 +23,7 @@ CalculateMeanOperation::CalculateMeanOperation()
void CalculateMeanOperation::init_execution()
{
image_reader_ = this->get_input_socket_reader(0);
is_calculated_ = false;
NodeOperation::init_mutex();
}
void CalculateMeanOperation::execute_pixel(float output[4], int /*x*/, int /*y*/, void * /*data*/)
{
output[0] = constant_value_;
}
void CalculateMeanOperation::deinit_execution()
{
image_reader_ = nullptr;
NodeOperation::deinit_mutex();
}
bool CalculateMeanOperation::determine_depending_area_of_interest(
rcti * /*input*/, ReadBufferOperation *read_operation, rcti *output)
{
rcti image_input;
if (is_calculated_) {
return false;
}
NodeOperation *operation = get_input_operation(0);
image_input.xmax = operation->get_width();
image_input.xmin = 0;
image_input.ymax = operation->get_height();
image_input.ymin = 0;
if (operation->determine_depending_area_of_interest(&image_input, read_operation, output)) {
return true;
}
return false;
}
void *CalculateMeanOperation::initialize_tile_data(rcti *rect)
{
lock_mutex();
if (!is_calculated_) {
MemoryBuffer *tile = (MemoryBuffer *)image_reader_->initialize_tile_data(rect);
constant_value_ = calculate_mean_tile(tile);
is_calculated_ = true;
}
unlock_mutex();
return nullptr;
}
float CalculateMeanOperation::calculate_mean_tile(MemoryBuffer *tile) const
{
float *buffer = tile->get_buffer();
int size = tile->get_width() * tile->get_height();
int pixels = 0;
float sum = 0.0f;
for (int i = 0, offset = 0; i < size; i++, offset += 4) {
if (buffer[offset + 3] > 0) {
pixels++;
switch (setting_) {
case 1: {
sum += IMB_colormanagement_get_luminance(&buffer[offset]);
break;
}
case 2: {
sum += buffer[offset];
break;
}
case 3: {
sum += buffer[offset + 1];
break;
}
case 4: {
sum += buffer[offset + 2];
break;
}
case 5: {
float yuv[3];
rgb_to_yuv(buffer[offset],
buffer[offset + 1],
buffer[offset + 2],
&yuv[0],
&yuv[1],
&yuv[2],
BLI_YUV_ITU_BT709);
sum += yuv[0];
break;
}
}
}
}
return sum / pixels;
}
void CalculateMeanOperation::set_setting(int setting)

24
source/blender/compositor/operations/COM_CalculateMeanOperation.h
View File

@ -22,11 +22,6 @@ class CalculateMeanOperation : public ConstantOperation {
};
protected:
/**
* \brief Cached reference to the reader
*/
SocketReader *image_reader_;
bool is_calculated_;
float constant_value_;
int setting_;
@ -35,26 +30,8 @@ class CalculateMeanOperation : public ConstantOperation {
public:
CalculateMeanOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel(float output[4], int x, int y, void *data) override;
/**
* Initialize the execution
*/
void init_execution() override;
void *initialize_tile_data(rcti *rect) override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
bool determine_depending_area_of_interest(rcti *input,
ReadBufferOperation *read_operation,
rcti *output) override;
void set_setting(int setting);
void get_area_of_interest(int input_idx, const rcti &output_area, rcti &r_input_area) override;
@ -71,7 +48,6 @@ class CalculateMeanOperation : public ConstantOperation {
* The caller takes care of checking the value is only calculated once. */
virtual float calculate_value(const MemoryBuffer *input) const;
float calculate_mean_tile(MemoryBuffer *tile) const;
float calculate_mean(const MemoryBuffer *input) const;
private:

71
source/blender/compositor/operations/COM_CalculateStandardDeviationOperation.cc
View File

@ -10,77 +10,6 @@
namespace blender::compositor {
void CalculateStandardDeviationOperation::execute_pixel(float output[4],
int /*x*/,
int /*y*/,
void * /*data*/)
{
output[0] = standard_deviation_;
}
void *CalculateStandardDeviationOperation::initialize_tile_data(rcti *rect)
{
lock_mutex();
if (!is_calculated_) {
MemoryBuffer *tile = (MemoryBuffer *)image_reader_->initialize_tile_data(rect);
standard_deviation_ = 0.0f;
float *buffer = tile->get_buffer();
int size = tile->get_width() * tile->get_height();
int pixels = 0;
float sum = 0.0f;
const float mean = this->calculate_mean_tile(tile);
for (int i = 0, offset = 0; i < size; i++, offset += 4) {
if (buffer[offset + 3] > 0) {
pixels++;
switch (setting_) {
case 1: /* rgb combined */
{
float value = IMB_colormanagement_get_luminance(&buffer[offset]);
sum += (value - mean) * (value - mean);
break;
}
case 2: /* red */
{
float value = buffer[offset];
sum += (value - mean) * (value - mean);
break;
}
case 3: /* green */
{
float value = buffer[offset + 1];
sum += (value - mean) * (value - mean);
break;
}
case 4: /* blue */
{
float value = buffer[offset + 2];
sum += (value - mean) * (value - mean);
break;
}
case 5: /* luminance */
{
float yuv[3];
rgb_to_yuv(buffer[offset],
buffer[offset + 1],
buffer[offset + 2],
&yuv[0],
&yuv[1],
&yuv[2],
BLI_YUV_ITU_BT709);
sum += (yuv[0] - mean) * (yuv[0] - mean);
break;
}
}
}
}
standard_deviation_ = sqrt(sum / float(pixels - 1));
is_calculated_ = true;
}
unlock_mutex();
return nullptr;
}
float CalculateStandardDeviationOperation::calculate_value(const MemoryBuffer *input) const
{
const float mean = this->calculate_mean(input);

9
source/blender/compositor/operations/COM_CalculateStandardDeviationOperation.h
View File

@ -19,15 +19,6 @@ class CalculateStandardDeviationOperation : public CalculateMeanOperation {
protected:
float standard_deviation_;
public:
/**
* The inner loop of this operation.
*/
void execute_pixel(float output[4], int x, int y, void *data) override;
void *initialize_tile_data(rcti *rect) override;
protected:
float calculate_value(const MemoryBuffer *input) const override;
private:

42
source/blender/compositor/operations/COM_ChangeHSVOperation.cc
View File

@ -13,51 +13,9 @@ ChangeHSVOperation::ChangeHSVOperation()
this->add_input_socket(DataType::Value);
this->add_input_socket(DataType::Value);
this->add_output_socket(DataType::Color);
input_operation_ = nullptr;
flags_.can_be_constant = true;
}
void ChangeHSVOperation::init_execution()
{
input_operation_ = get_input_socket_reader(0);
hue_operation_ = get_input_socket_reader(1);
saturation_operation_ = get_input_socket_reader(2);
value_operation_ = get_input_socket_reader(3);
}
void ChangeHSVOperation::deinit_execution()
{
input_operation_ = nullptr;
hue_operation_ = nullptr;
saturation_operation_ = nullptr;
value_operation_ = nullptr;
}
void ChangeHSVOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_color1[4];
float hue[4], saturation[4], value[4];
input_operation_->read_sampled(input_color1, x, y, sampler);
hue_operation_->read_sampled(hue, x, y, sampler);
saturation_operation_->read_sampled(saturation, x, y, sampler);
value_operation_->read_sampled(value, x, y, sampler);
output[0] = input_color1[0] + (hue[0] - 0.5f);
if (output[0] > 1.0f) {
output[0] -= 1.0f;
}
else if (output[0] < 0.0f) {
output[0] += 1.0f;
}
output[1] = input_color1[1] * saturation[0];
output[2] = input_color1[2] * value[0];
output[3] = input_color1[3];
}
void ChangeHSVOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)

17
source/blender/compositor/operations/COM_ChangeHSVOperation.h
View File

@ -13,26 +13,9 @@ namespace blender::compositor {
* it assumes we are in sRGB color space.
*/
class ChangeHSVOperation : public MultiThreadedOperation {
private:
SocketReader *input_operation_;
SocketReader *hue_operation_;
SocketReader *saturation_operation_;
SocketReader *value_operation_;
public:
/**
* Default constructor
*/
ChangeHSVOperation();
void init_execution() override;
void deinit_execution() override;
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs) override;

47
source/blender/compositor/operations/COM_ChannelMatteOperation.cc
View File

@ -11,14 +11,11 @@ ChannelMatteOperation::ChannelMatteOperation()
add_input_socket(DataType::Color);
add_output_socket(DataType::Value);
input_image_program_ = nullptr;
flags_.can_be_constant = true;
}
void ChannelMatteOperation::init_execution()
{
input_image_program_ = this->get_input_socket_reader(0);
limit_range_ = limit_max_ - limit_min_;
switch (limit_method_) {
@ -63,50 +60,6 @@ void ChannelMatteOperation::init_execution()
}
}
void ChannelMatteOperation::deinit_execution()
{
input_image_program_ = nullptr;
}
void ChannelMatteOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float in_color[4];
float alpha;
const float limit_max = limit_max_;
const float limit_min = limit_min_;
const float limit_range = limit_range_;
input_image_program_->read_sampled(in_color, x, y, sampler);
/* matte operation */
alpha = in_color[ids_[0]] - std::max(in_color[ids_[1]], in_color[ids_[2]]);
/* flip because 0.0 is transparent, not 1.0 */
alpha = 1.0f - alpha;
/* test range */
if (alpha > limit_max) {
alpha = in_color[3]; /* Whatever it was prior. */
}
else if (alpha < limit_min) {
alpha = 0.0f;
}
else { /* Blend. */
alpha = (alpha - limit_min) / limit_range;
}
/* Store matte(alpha) value in [0] to go with
* COM_SetAlphaMultiplyOperation and the Value output.
*/
/* Don't make something that was more transparent less transparent. */
output[0] = std::min(alpha, in_color[3]);
}
void ChannelMatteOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)

11
source/blender/compositor/operations/COM_ChannelMatteOperation.h
View File

@ -14,8 +14,6 @@ namespace blender::compositor {
*/
class ChannelMatteOperation : public MultiThreadedOperation {
private:
SocketReader *input_image_program_;
// int color_space_; /* node->custom1 */ /* UNUSED */ /* TODO? */
int matte_channel_; /* node->custom2 */
int limit_method_; /* node->algorithm */
@ -37,18 +35,9 @@ class ChannelMatteOperation : public MultiThreadedOperation {
int ids_[3];
public:
/**
* Default constructor
*/
ChannelMatteOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void init_execution() override;
void deinit_execution() override;
void set_settings(NodeChroma *node_chroma, const int custom2)
{

81
source/blender/compositor/operations/COM_ChromaMatteOperation.cc
View File

@ -12,90 +12,9 @@ ChromaMatteOperation::ChromaMatteOperation()
add_input_socket(DataType::Color);
add_output_socket(DataType::Value);
input_image_program_ = nullptr;
input_key_program_ = nullptr;
flags_.can_be_constant = true;
}
void ChromaMatteOperation::init_execution()
{
input_image_program_ = this->get_input_socket_reader(0);
input_key_program_ = this->get_input_socket_reader(1);
}
void ChromaMatteOperation::deinit_execution()
{
input_image_program_ = nullptr;
input_key_program_ = nullptr;
}
void ChromaMatteOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float in_key[4];
float in_image[4];
const float acceptance = settings_->t1; /* in radians */
const float cutoff = settings_->t2; /* in radians */
const float gain = settings_->fstrength;
float x_angle, z_angle, alpha;
float theta, beta;
float kfg;
input_key_program_->read_sampled(in_key, x, y, sampler);
input_image_program_->read_sampled(in_image, x, y, sampler);
/* Store matte(alpha) value in [0] to go with
* #COM_SetAlphaMultiplyOperation and the Value output. */
/* Algorithm from book "Video Demystified", does not include the spill reduction part. */
/* Find theta, the angle that the color space should be rotated based on key. */
/* rescale to -1.0..1.0 */
// in_image[0] = (in_image[0] * 2.0f) - 1.0f; // UNUSED
in_image[1] = (in_image[1] * 2.0f) - 1.0f;
in_image[2] = (in_image[2] * 2.0f) - 1.0f;
// in_key[0] = (in_key[0] * 2.0f) - 1.0f; // UNUSED
in_key[1] = (in_key[1] * 2.0f) - 1.0f;
in_key[2] = (in_key[2] * 2.0f) - 1.0f;
theta = atan2(in_key[2], in_key[1]);
/* Rotate the cb and cr into x/z space. */
x_angle = in_image[1] * cosf(theta) + in_image[2] * sinf(theta);
z_angle = in_image[2] * cosf(theta) - in_image[1] * sinf(theta);
/* If within the acceptance angle. */
/* If kfg is <0 then the pixel is outside of the key color. */
kfg = x_angle - (fabsf(z_angle) / tanf(acceptance / 2.0f));
if (kfg > 0.0f) { /* found a pixel that is within key color */
alpha = 1.0f - (kfg / gain);
beta = atan2(z_angle, x_angle);
/* if beta is within the cutoff angle */
if (fabsf(beta) < (cutoff / 2.0f)) {
alpha = 0.0f;
}
/* don't make something that was more transparent less transparent */
if (alpha < in_image[3]) {
output[0] = alpha;
}
else {
output[0] = in_image[3];
}
}
else { /* Pixel is outside key color. */
output[0] = in_image[3]; /* Make pixel just as transparent as it was before. */
}
}
void ChromaMatteOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)

13
source/blender/compositor/operations/COM_ChromaMatteOperation.h
View File

@ -15,23 +15,10 @@ namespace blender::compositor {
class ChromaMatteOperation : public MultiThreadedOperation {
private:
NodeChroma *settings_;
SocketReader *input_image_program_;
SocketReader *input_key_program_;
public:
/**
* Default constructor
*/
ChromaMatteOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
void init_execution() override;
void deinit_execution() override;
void set_settings(NodeChroma *node_chroma)
{
settings_ = node_chroma;

38
source/blender/compositor/operations/COM_ColorBalanceASCCDLOperation.cc
View File

@ -23,42 +23,10 @@ ColorBalanceASCCDLOperation::ColorBalanceASCCDLOperation()
this->add_input_socket(DataType::Value);
this->add_input_socket(DataType::Color);
this->add_output_socket(DataType::Color);
input_value_operation_ = nullptr;
input_color_operation_ = nullptr;
this->set_canvas_input_index(1);
flags_.can_be_constant = true;
}
void ColorBalanceASCCDLOperation::init_execution()
{
input_value_operation_ = this->get_input_socket_reader(0);
input_color_operation_ = this->get_input_socket_reader(1);
}
void ColorBalanceASCCDLOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_color[4];
float value[4];
input_value_operation_->read_sampled(value, x, y, sampler);
input_color_operation_->read_sampled(input_color, x, y, sampler);
float fac = value[0];
fac = std::min(1.0f, fac);
const float mfac = 1.0f - fac;
output[0] = mfac * input_color[0] +
fac * colorbalance_cdl(input_color[0], offset_[0], power_[0], slope_[0]);
output[1] = mfac * input_color[1] +
fac * colorbalance_cdl(input_color[1], offset_[1], power_[1], slope_[1]);
output[2] = mfac * input_color[2] +
fac * colorbalance_cdl(input_color[2], offset_[2], power_[2], slope_[2]);
output[3] = input_color[3];
}
void ColorBalanceASCCDLOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {
@ -76,10 +44,4 @@ void ColorBalanceASCCDLOperation::update_memory_buffer_row(PixelCursor &p)
}
}
void ColorBalanceASCCDLOperation::deinit_execution()
{
input_value_operation_ = nullptr;
input_color_operation_ = nullptr;
}
} // namespace blender::compositor

24
source/blender/compositor/operations/COM_ColorBalanceASCCDLOperation.h
View File

@ -14,37 +14,13 @@ namespace blender::compositor {
*/
class ColorBalanceASCCDLOperation : public MultiThreadedRowOperation {
protected:
/**
* Prefetched reference to the input_program
*/
SocketReader *input_value_operation_;
SocketReader *input_color_operation_;
float offset_[3];
float power_[3];
float slope_[3];
public:
/**
* Default constructor
*/
ColorBalanceASCCDLOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_offset(float offset[3])
{
copy_v3_v3(offset_, offset);

38
source/blender/compositor/operations/COM_ColorBalanceLGGOperation.cc
View File

@ -30,42 +30,10 @@ ColorBalanceLGGOperation::ColorBalanceLGGOperation()
this->add_input_socket(DataType::Value);
this->add_input_socket(DataType::Color);
this->add_output_socket(DataType::Color);
input_value_operation_ = nullptr;
input_color_operation_ = nullptr;
this->set_canvas_input_index(1);
flags_.can_be_constant = true;
}
void ColorBalanceLGGOperation::init_execution()
{
input_value_operation_ = this->get_input_socket_reader(0);
input_color_operation_ = this->get_input_socket_reader(1);
}
void ColorBalanceLGGOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_color[4];
float value[4];
input_value_operation_->read_sampled(value, x, y, sampler);
input_color_operation_->read_sampled(input_color, x, y, sampler);
float fac = value[0];
fac = std::min(1.0f, fac);
const float mfac = 1.0f - fac;
output[0] = mfac * input_color[0] +
fac * colorbalance_lgg(input_color[0], lift_[0], gamma_inv_[0], gain_[0]);
output[1] = mfac * input_color[1] +
fac * colorbalance_lgg(input_color[1], lift_[1], gamma_inv_[1], gain_[1]);
output[2] = mfac * input_color[2] +
fac * colorbalance_lgg(input_color[2], lift_[2], gamma_inv_[2], gain_[2]);
output[3] = input_color[3];
}
void ColorBalanceLGGOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {
@ -83,10 +51,4 @@ void ColorBalanceLGGOperation::update_memory_buffer_row(PixelCursor &p)
}
}
void ColorBalanceLGGOperation::deinit_execution()
{
input_value_operation_ = nullptr;
input_color_operation_ = nullptr;
}
} // namespace blender::compositor

24
source/blender/compositor/operations/COM_ColorBalanceLGGOperation.h
View File

@ -14,37 +14,13 @@ namespace blender::compositor {
*/
class ColorBalanceLGGOperation : public MultiThreadedRowOperation {
protected:
/**
* Prefetched reference to the input_program
*/
SocketReader *input_value_operation_;
SocketReader *input_color_operation_;
float gain_[3];
float lift_[3];
float gamma_inv_[3];
public:
/**
* Default constructor
*/
ColorBalanceLGGOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_gain(const float gain[3])
{
copy_v3_v3(gain_, gain);

118
source/blender/compositor/operations/COM_ColorCorrectionOperation.cc
View File

@ -13,18 +13,11 @@ ColorCorrectionOperation::ColorCorrectionOperation()
this->add_input_socket(DataType::Color);
this->add_input_socket(DataType::Value);
this->add_output_socket(DataType::Color);
input_image_ = nullptr;
input_mask_ = nullptr;
red_channel_enabled_ = true;
green_channel_enabled_ = true;
blue_channel_enabled_ = true;
flags_.can_be_constant = true;
}
void ColorCorrectionOperation::init_execution()
{
input_image_ = this->get_input_socket_reader(0);
input_mask_ = this->get_input_socket_reader(1);
}
/* Calculate x^y if the function is defined. Otherwise return the given fallback value. */
BLI_INLINE float color_correct_powf_safe(const float x, const float y, const float fallback_value)
@ -35,111 +28,6 @@ BLI_INLINE float color_correct_powf_safe(const float x, const float y, const flo
return powf(x, y);
}
void ColorCorrectionOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_image_color[4];
float input_mask[4];
input_image_->read_sampled(input_image_color, x, y, sampler);
input_mask_->read_sampled(input_mask, x, y, sampler);
float level = (input_image_color[0] + input_image_color[1] + input_image_color[2]) / 3.0f;
float contrast = data_->master.contrast;
float saturation = data_->master.saturation;
float gamma = data_->master.gamma;
float gain = data_->master.gain;
float lift = data_->master.lift;
float r, g, b;
float value = input_mask[0];
value = std::min(1.0f, value);
const float mvalue = 1.0f - value;
float level_shadows = 0.0;
float level_midtones = 0.0;
float level_highlights = 0.0;
#define MARGIN 0.10f
#define MARGIN_DIV (0.5f / MARGIN)
if (level < data_->startmidtones - MARGIN) {
level_shadows = 1.0f;
}
else if (level < data_->startmidtones + MARGIN) {
level_midtones = ((level - data_->startmidtones) * MARGIN_DIV) + 0.5f;
level_shadows = 1.0f - level_midtones;
}
else if (level < data_->endmidtones - MARGIN) {
level_midtones = 1.0f;
}
else if (level < data_->endmidtones + MARGIN) {
level_highlights = ((level - data_->endmidtones) * MARGIN_DIV) + 0.5f;
level_midtones = 1.0f - level_highlights;
}
else {
level_highlights = 1.0f;
}
#undef MARGIN
#undef MARGIN_DIV
contrast *= (level_shadows * data_->shadows.contrast) +
(level_midtones * data_->midtones.contrast) +
(level_highlights * data_->highlights.contrast);
saturation *= (level_shadows * data_->shadows.saturation) +
(level_midtones * data_->midtones.saturation) +
(level_highlights * data_->highlights.saturation);
gamma *= (level_shadows * data_->shadows.gamma) + (level_midtones * data_->midtones.gamma) +
(level_highlights * data_->highlights.gamma);
gain *= (level_shadows * data_->shadows.gain) + (level_midtones * data_->midtones.gain) +
(level_highlights * data_->highlights.gain);
lift += (level_shadows * data_->shadows.lift) + (level_midtones * data_->midtones.lift) +
(level_highlights * data_->highlights.lift);
float invgamma = 1.0f / gamma;
float luma = IMB_colormanagement_get_luminance(input_image_color);
r = input_image_color[0];
g = input_image_color[1];
b = input_image_color[2];
r = (luma + saturation * (r - luma));
g = (luma + saturation * (g - luma));
b = (luma + saturation * (b - luma));
r = 0.5f + ((r - 0.5f) * contrast);
g = 0.5f + ((g - 0.5f) * contrast);
b = 0.5f + ((b - 0.5f) * contrast);
/* Check for negative values to avoid nan. */
r = color_correct_powf_safe(r * gain + lift, invgamma, r);
g = color_correct_powf_safe(g * gain + lift, invgamma, g);
b = color_correct_powf_safe(b * gain + lift, invgamma, b);
/* Mix with mask. */
r = mvalue * input_image_color[0] + value * r;
g = mvalue * input_image_color[1] + value * g;
b = mvalue * input_image_color[2] + value * b;
if (red_channel_enabled_) {
output[0] = r;
}
else {
output[0] = input_image_color[0];
}
if (green_channel_enabled_) {
output[1] = g;
}
else {
output[1] = input_image_color[1];
}
if (blue_channel_enabled_) {
output[2] = b;
}
else {
output[2] = input_image_color[2];
}
output[3] = input_image_color[3];
}
void ColorCorrectionOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {
@ -217,10 +105,4 @@ void ColorCorrectionOperation::update_memory_buffer_row(PixelCursor &p)
}
}
void ColorCorrectionOperation::deinit_execution()
{
input_image_ = nullptr;
input_mask_ = nullptr;
}
} // namespace blender::compositor

20
source/blender/compositor/operations/COM_ColorCorrectionOperation.h
View File

@ -10,11 +10,6 @@ namespace blender::compositor {
class ColorCorrectionOperation : public MultiThreadedRowOperation {
private:
/**
* Cached reference to the input_program
*/
SocketReader *input_image_;
SocketReader *input_mask_;
NodeColorCorrection *data_;
bool red_channel_enabled_;
@ -24,21 +19,6 @@ class ColorCorrectionOperation : public MultiThreadedRowOperation {
public:
ColorCorrectionOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_data(NodeColorCorrection *data)
{
data_ = data;

94
source/blender/compositor/operations/COM_ColorCurveOperation.cc
View File

@ -16,72 +16,15 @@ ColorCurveOperation::ColorCurveOperation()
this->add_input_socket(DataType::Color);
this->add_output_socket(DataType::Color);
input_fac_program_ = nullptr;
input_image_program_ = nullptr;
input_black_program_ = nullptr;
input_white_program_ = nullptr;
this->set_canvas_input_index(1);
}
void ColorCurveOperation::init_execution()
{
CurveBaseOperation::init_execution();
input_fac_program_ = this->get_input_socket_reader(0);
input_image_program_ = this->get_input_socket_reader(1);
input_black_program_ = this->get_input_socket_reader(2);
input_white_program_ = this->get_input_socket_reader(3);
BKE_curvemapping_premultiply(curve_mapping_, false);
}
void ColorCurveOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
CurveMapping *cumap = curve_mapping_;
float fac[4];
float image[4];
/* local versions of cumap->black, cumap->white, cumap->bwmul */
float black[4];
float white[4];
float bwmul[3];
input_black_program_->read_sampled(black, x, y, sampler);
input_white_program_->read_sampled(white, x, y, sampler);
/* get our own local bwmul value,
* since we can't be threadsafe and use cumap->bwmul & friends */
BKE_curvemapping_set_black_white_ex(black, white, bwmul);
input_fac_program_->read_sampled(fac, x, y, sampler);
input_image_program_->read_sampled(image, x, y, sampler);
if (*fac >= 1.0f) {
BKE_curvemapping_evaluate_premulRGBF_ex(cumap, output, image, black, bwmul);
}
else if (*fac <= 0.0f) {
copy_v3_v3(output, image);
}
else {
float col[4];
BKE_curvemapping_evaluate_premulRGBF_ex(cumap, col, image, black, bwmul);
interp_v3_v3v3(output, image, col, *fac);
}
output[3] = image[3];
}
void ColorCurveOperation::deinit_execution()
{
CurveBaseOperation::deinit_execution();
input_fac_program_ = nullptr;
input_image_program_ = nullptr;
input_black_program_ = nullptr;
input_white_program_ = nullptr;
}
void ColorCurveOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)
@ -120,54 +63,17 @@ ConstantLevelColorCurveOperation::ConstantLevelColorCurveOperation()
this->add_input_socket(DataType::Color);
this->add_output_socket(DataType::Color);
input_fac_program_ = nullptr;
input_image_program_ = nullptr;
this->set_canvas_input_index(1);
}
void ConstantLevelColorCurveOperation::init_execution()
{
CurveBaseOperation::init_execution();
input_fac_program_ = this->get_input_socket_reader(0);
input_image_program_ = this->get_input_socket_reader(1);
BKE_curvemapping_premultiply(curve_mapping_, false);
BKE_curvemapping_set_black_white(curve_mapping_, black_, white_);
}
void ConstantLevelColorCurveOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float fac[4];
float image[4];
input_fac_program_->read_sampled(fac, x, y, sampler);
input_image_program_->read_sampled(image, x, y, sampler);
if (*fac >= 1.0f) {
BKE_curvemapping_evaluate_premulRGBF(curve_mapping_, output, image);
}
else if (*fac <= 0.0f) {
copy_v3_v3(output, image);
}
else {
float col[4];
BKE_curvemapping_evaluate_premulRGBF(curve_mapping_, col, image);
interp_v3_v3v3(output, image, col, *fac);
}
output[3] = image[3];
}
void ConstantLevelColorCurveOperation::deinit_execution()
{
CurveBaseOperation::deinit_execution();
input_fac_program_ = nullptr;
input_image_program_ = nullptr;
}
void ConstantLevelColorCurveOperation::update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs)

40
source/blender/compositor/operations/COM_ColorCurveOperation.h
View File

@ -9,33 +9,11 @@
namespace blender::compositor {
class ColorCurveOperation : public CurveBaseOperation {
private:
/**
* Cached reference to the input_program
*/
SocketReader *input_fac_program_;
SocketReader *input_image_program_;
SocketReader *input_black_program_;
SocketReader *input_white_program_;
public:
ColorCurveOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void update_memory_buffer_partial(MemoryBuffer *output,
const rcti &area,
Span<MemoryBuffer *> inputs) override;
@ -43,32 +21,14 @@ class ColorCurveOperation : public CurveBaseOperation {
class ConstantLevelColorCurveOperation : public CurveBaseOperation {
private:
/**
* Cached reference to the input_program
*/
SocketReader *input_fac_program_;
SocketReader *input_image_program_;
float black_[3];
float white_[3];
public:
ConstantLevelColorCurveOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void set_black_level(float black[3])
{
copy_v3_v3(black_, black);

31
source/blender/compositor/operations/COM_ColorExposureOperation.cc
View File

@ -11,34 +11,9 @@ ExposureOperation::ExposureOperation()
this->add_input_socket(DataType::Color);
this->add_input_socket(DataType::Value);
this->add_output_socket(DataType::Color);
input_program_ = nullptr;
flags_.can_be_constant = true;
}
void ExposureOperation::init_execution()
{
input_program_ = this->get_input_socket_reader(0);
input_exposure_program_ = this->get_input_socket_reader(1);
}
void ExposureOperation::execute_pixel_sampled(float output[4],
float x,
float y,
PixelSampler sampler)
{
float input_value[4];
float input_exposure[4];
input_program_->read_sampled(input_value, x, y, sampler);
input_exposure_program_->read_sampled(input_exposure, x, y, sampler);
const float exposure = pow(2, input_exposure[0]);
output[0] = input_value[0] * exposure;
output[1] = input_value[1] * exposure;
output[2] = input_value[2] * exposure;
output[3] = input_value[3];
}
void ExposureOperation::update_memory_buffer_row(PixelCursor &p)
{
for (; p.out < p.row_end; p.next()) {
@ -52,10 +27,4 @@ void ExposureOperation::update_memory_buffer_row(PixelCursor &p)
}
}
void ExposureOperation::deinit_execution()
{
input_program_ = nullptr;
input_exposure_program_ = nullptr;
}
} // namespace blender::compositor

22
source/blender/compositor/operations/COM_ColorExposureOperation.h
View File

@ -9,31 +9,9 @@
namespace blender::compositor {
class ExposureOperation : public MultiThreadedRowOperation {
private:
/**
* Cached reference to the input_program
*/
SocketReader *input_program_;
SocketReader *input_exposure_program_;
public:
ExposureOperation();
/**
* The inner loop of this operation.
*/
void execute_pixel_sampled(float output[4], float x, float y, PixelSampler sampler) override;
/**
* Initialize the execution
*/
void init_execution() override;
/**
* Deinitialize the execution
*/
void deinit_execution() override;
void update_memory_buffer_row(PixelCursor &p) override;
};

Some files were not shown because too many files have changed in this diff Show More