752 changed files with 11034 additions and 9747 deletions
--- a/build_files/build_environment/patches/cmake/modules/FindIlmBase.cmake
+++ b/build_files/build_environment/patches/cmake/modules/FindIlmBase.cmake
@ -8,7 +8,7 @@
 # It also supports non-standard names for the library components.
 #
 # To use a custom IlmBase:
-#   - Set the variable ILMBASE_CUSTOM to True
+#   - Set the variable ILMBASE_CUSTOM to TRUE
 #   - Set the variable ILMBASE_CUSTOM_LIBRARIES to a list of the libraries to
 #     use, e.g. "SpiImath SpiHalf SpiIlmThread SpiIex"
 #   - Optionally set the variable ILMBASE_CUSTOM_INCLUDE_DIR to any
@ -20,7 +20,7 @@
 #
 # ILMBASE_INCLUDE_DIR - where to find half.h, IlmBaseConfig.h, etc.
 # ILMBASE_LIBRARIES   - list of libraries to link against when using IlmBase.
-# ILMBASE_FOUND       - True if IlmBase was found.
+# ILMBASE_FOUND       - TRUE if IlmBase was found.

 # Other standard issue macros
 include(FindPackageHandleStandardArgs)
--- a/build_files/build_environment/patches/cmake/modules/FindOpenEXR.cmake
+++ b/build_files/build_environment/patches/cmake/modules/FindOpenEXR.cmake
@ -8,7 +8,7 @@
 # It also supports non-standard names for the library components.
 #
 # To use a custom OpenEXR
-#   - Set the variable OPENEXR_CUSTOM to True
+#   - Set the variable OPENEXR_CUSTOM to TRUE
 #   - Set the variable OPENEXR_CUSTOM_LIBRARY to the name of the library to
 #     use, e.g. "SpiIlmImf"
 #   - Optionally set the variable OPENEXR_CUSTOM_INCLUDE_DIR to any
@ -22,7 +22,7 @@
 # OPENEXR_LIBRARIES   - list of libraries to link against when using OpenEXR.
 #                       This list does NOT include the IlmBase libraries.
 #                       These are defined by the FindIlmBase module.
-# OPENEXR_FOUND       - True if OpenEXR was found.
+# OPENEXR_FOUND       - TRUE if OpenEXR was found.

 # Other standard issue macros
 include(SelectLibraryConfigurations)
--- a/build_files/cmake/platform/platform_unix.cmake
+++ b/build_files/cmake/platform/platform_unix.cmake
@ -20,14 +20,14 @@ else()
    # Choose the best suitable libraries.
    if(EXISTS ${LIBDIR_NATIVE_ABI})
      set(LIBDIR ${LIBDIR_NATIVE_ABI})
-      set(WITH_LIBC_MALLOC_HOOK_WORKAROUND True)
+      set(WITH_LIBC_MALLOC_HOOK_WORKAROUND TRUE)
    elseif(EXISTS ${LIBDIR_GLIBC228_ABI})
      set(LIBDIR ${LIBDIR_GLIBC228_ABI})
      if(WITH_MEM_JEMALLOC)
        # jemalloc provides malloc hooks.
-        set(WITH_LIBC_MALLOC_HOOK_WORKAROUND False)
+        set(WITH_LIBC_MALLOC_HOOK_WORKAROUND FALSE)
      else()
-        set(WITH_LIBC_MALLOC_HOOK_WORKAROUND True)
+        set(WITH_LIBC_MALLOC_HOOK_WORKAROUND TRUE)
      endif()
    endif()

--- a/build_files/cmake/platform/platform_win32.cmake
+++ b/build_files/cmake/platform/platform_win32.cmake
@ -39,8 +39,14 @@ if(CMAKE_C_COMPILER_ID MATCHES "Clang")
    set(WITH_WINDOWS_STRIPPED_PDB OFF)
  endif()
 else()
-  if(WITH_BLENDER AND CMAKE_CXX_COMPILER_VERSION VERSION_LESS 19.28.29921) # MSVC 2019 16.9.16
-    message(FATAL_ERROR "Compiler is unsupported, MSVC 2019 16.9.16 or newer is required for building blender.")
+  if(WITH_BLENDER)
+    if(CMAKE_CXX_COMPILER_VERSION VERSION_LESS 19.28.29921) # MSVC 2019 16.9.16
+      message(FATAL_ERROR "Compiler is unsupported, MSVC 2019 16.9.16 or newer is required for building blender.")
+    endif()
+    if(CMAKE_CXX_COMPILER_VERSION VERSION_GREATER_EQUAL 19.36.32532 AND # MSVC 2022 17.6.0 has a bad codegen
+       CMAKE_CXX_COMPILER_VERSION VERSION_LESS 19.37.32705)             # But it is fixed in 2022 17.7 preview 1
+      message(FATAL_ERROR "Compiler is unsupported, MSVC 2022 17.6.x has codegen issues and cannot be used to build blender. Please use MSVC 17.5 for the time being.")
+    endif()
  endif()
 endif()

@ -1034,7 +1040,7 @@ endif()

 if(WITH_VULKAN_BACKEND)
  if(EXISTS ${LIBDIR}/vulkan)
-    set(VULKAN_FOUND On)
+    set(VULKAN_FOUND ON)
    set(VULKAN_ROOT_DIR ${LIBDIR}/vulkan)
    set(VULKAN_INCLUDE_DIR ${VULKAN_ROOT_DIR}/include)
    set(VULKAN_INCLUDE_DIRS ${VULKAN_INCLUDE_DIR})
@ -1048,7 +1054,7 @@ endif()

 if(WITH_VULKAN_BACKEND)
  if(EXISTS ${LIBDIR}/shaderc)
-    set(SHADERC_FOUND On)
+    set(SHADERC_FOUND ON)
    set(SHADERC_ROOT_DIR ${LIBDIR}/shaderc)
    set(SHADERC_INCLUDE_DIR ${SHADERC_ROOT_DIR}/include)
    set(SHADERC_INCLUDE_DIRS ${SHADERC_INCLUDE_DIR})
--- a/doc/doxygen/Doxyfile
+++ b/doc/doxygen/Doxyfile
@ -38,7 +38,7 @@ PROJECT_NAME           = Blender
 # could be handy for archiving the generated documentation or if some version
 # control system is used.

-PROJECT_NUMBER         = V3.6
+PROJECT_NUMBER         = V4.0

 # Using the PROJECT_BRIEF tag one can provide an optional one line description
 # for a project that appears at the top of each page and should give viewer a
--- a/extern/glog/README.blender
+++ b/extern/glog/README.blender
@ -8,3 +8,5 @@ Local modifications:
 * Added special definitions of HAVE_SNPRINTF and HAVE_LIB_GFLAGS
  in Windows' specific config.h.
 * Silenced syscall deprecation warnings on macOS >= 10.12.
+* Usage of syscall() is not allowed and use getthrid() to
+  retreive the thread ID on OpenBSD
--- a/extern/glog/src/config.h
+++ b/extern/glog/src/config.h
@ -2,6 +2,8 @@
 #include "config_mac.h"
 #elif defined(__FreeBSD__) || defined(__FreeBSD_kernel__)
 #include "config_freebsd.h"
+#elif defined(__OpenBSD__)
+ #include "config_openbsd.h"
 #elif defined(__MINGW32__)
 #include "windows/config.h"
 #elif defined(__linux__)
--- a/extern/glog/src/config_openbsd.h
+++ b/extern/glog/src/config_openbsd.h
@ -0,0 +1,192 @@
+/* define if glog doesn't use RTTI */
+/* #undef DISABLE_RTTI */
+
+/* Namespace for Google classes */
+#define GOOGLE_NAMESPACE google
+
+/* Define if you have the `dladdr' function */
+/* #undef HAVE_DLADDR */
+
+/* Define if you have the `snprintf' function */
+#define HAVE_SNPRINTF
+
+/* Define to 1 if you have the <dlfcn.h> header file. */
+#define HAVE_DLFCN_H
+
+/* Define to 1 if you have the <execinfo.h> header file. */
+/* #undef HAVE_EXECINFO_H */
+
+/* Define if you have the `fcntl' function */
+#define HAVE_FCNTL
+
+/* Define to 1 if you have the <glob.h> header file. */
+#define HAVE_GLOB_H
+
+/* Define to 1 if you have the <inttypes.h> header file. */
+#define HAVE_INTTYPES_H 1
+
+/* Define to 1 if you have the `pthread' library (-lpthread). */
+#define HAVE_LIBPTHREAD
+
+/* Define to 1 if you have the <libunwind.h> header file. */
+/* #undef HAVE_LIBUNWIND_H */
+
+/* define if you have google gflags library */
+#define HAVE_LIB_GFLAGS
+
+/* define if you have google gmock library */
+/* #undef HAVE_LIB_GMOCK */
+
+/* define if you have google gtest library */
+/* #undef HAVE_LIB_GTEST */
+
+/* define if you have libunwind */
+/* #undef HAVE_LIB_UNWIND */
+
+/* Define to 1 if you have the <memory.h> header file. */
+#define HAVE_MEMORY_H
+
+/* define to disable multithreading support. */
+/* #undef NO_THREADS */
+
+/* define if the compiler implements namespaces */
+#define HAVE_NAMESPACES
+
+/* Define if you have the 'pread' function */
+#define HAVE_PREAD
+
+/* Define if you have POSIX threads libraries and header files. */
+#define HAVE_PTHREAD
+
+/* Define to 1 if you have the <pwd.h> header file. */
+#define HAVE_PWD_H
+
+/* Define if you have the 'pwrite' function */
+#define HAVE_PWRITE
+
+/* define if the compiler implements pthread_rwlock_* */
+#define HAVE_RWLOCK 1
+
+/* Define if you have the 'sigaction' function */
+#define HAVE_SIGACTION
+
+/* Define if you have the `sigaltstack' function */
+#define HAVE_SIGALTSTACK 1
+
+/* Define to 1 if you have the <stdint.h> header file. */
+#define HAVE_STDINT_H 1
+
+/* Define to 1 if you have the <stdlib.h> header file. */
+#define HAVE_STDLIB_H
+
+/* Define to 1 if you have the <strings.h> header file. */
+#define HAVE_STRINGS_H
+
+/* Define to 1 if you have the <string.h> header file. */
+#define HAVE_STRING_H
+
+/* Define to 1 if you have the <syscall.h> header file. */
+/* #undef HAVE_SYSCALL_H */
+
+/* Define to 1 if you have the <syslog.h> header file. */
+#define HAVE_SYSLOG_H
+
+/* Define to 1 if you have the <sys/stat.h> header file. */
+#define HAVE_SYS_STAT_H
+
+/* Define to 1 if you have the <sys/syscall.h> header file. */
+#define HAVE_SYS_SYSCALL_H
+
+/* Define to 1 if you have the <sys/time.h> header file. */
+#define HAVE_SYS_TIME_H
+
+/* Define to 1 if you have the <sys/types.h> header file. */
+#define HAVE_SYS_TYPES_H 1
+
+/* Define to 1 if you have the <sys/ucontext.h> header file. */
+/* #undef HAVE_SYS_UCONTEXT_H */
+
+/* Define to 1 if you have the <sys/utsname.h> header file. */
+#define HAVE_SYS_UTSNAME_H
+
+/* Define to 1 if you have the <ucontext.h> header file. */
+/* #undef HAVE_UCONTEXT_H */
+
+/* Define to 1 if you have the <unistd.h> header file. */
+#define HAVE_UNISTD_H 1
+
+/* Define to 1 if you have the <unwind.h> header file. */
+#define HAVE_UNWIND_H 1
+
+/* define if the compiler supports using expression for operator */
+#define HAVE_USING_OPERATOR
+
+/* define if your compiler has __attribute__ */
+#define HAVE___ATTRIBUTE__
+
+/* define if your compiler has __builtin_expect */
+#define HAVE___BUILTIN_EXPECT 1
+
+/* define if your compiler has __sync_val_compare_and_swap */
+#define HAVE___SYNC_VAL_COMPARE_AND_SWAP
+
+/* Define to the sub-directory in which libtool stores uninstalled libraries.
+   */
+/* #undef LT_OBJDIR */
+
+/* Name of package */
+/* #undef PACKAGE */
+
+/* Define to the address where bug reports for this package should be sent. */
+/* #undef PACKAGE_BUGREPORT */
+
+/* Define to the full name of this package. */
+/* #undef PACKAGE_NAME */
+
+/* Define to the full name and version of this package. */
+/* #undef PACKAGE_STRING */
+
+/* Define to the one symbol short name of this package. */
+/* #undef PACKAGE_TARNAME */
+
+/* Define to the home page for this package. */
+/* #undef PACKAGE_URL */
+
+/* Define to the version of this package. */
+/* #undef PACKAGE_VERSION */
+
+/* How to access the PC from a struct ucontext */
+/* #undef PC_FROM_UCONTEXT */
+
+/* Define to necessary symbol if this constant uses a non-standard name on
+   your system. */
+/* #undef PTHREAD_CREATE_JOINABLE */
+
+/* The size of `void *', as computed by sizeof. */
+#define SIZEOF_VOID_P 8
+
+/* Define to 1 if you have the ANSI C header files. */
+#define STDC_HEADERS 1
+
+/* the namespace where STL code like vector<> is defined */
+#define STL_NAMESPACE std
+
+/* location of source code */
+#define TEST_SRC_DIR "."
+
+/* Version number of package */
+/* #undef VERSION */
+
+/* Stops putting the code inside the Google namespace */
+#define _END_GOOGLE_NAMESPACE_ }
+
+/* Puts following code inside the Google namespace */
+#define _START_GOOGLE_NAMESPACE_ namespace google {
+
+#define GOOGLE_GLOG_DLL_DECL
+
+/* isn't getting defined by configure script when clang compilers are used
+   and cuases compilation errors in stactrace/unwind modules */
+#ifdef __clang__
+#  define NO_FRAME_POINTER
+#endif
--- a/extern/glog/src/raw_logging.cc
+++ b/extern/glog/src/raw_logging.cc
@ -59,7 +59,8 @@
 # include <unistd.h>
 #endif

-#if (defined(HAVE_SYSCALL_H) || defined(HAVE_SYS_SYSCALL_H)) && (!(defined OS_MACOSX))
+#if (defined(HAVE_SYSCALL_H) || defined(HAVE_SYS_SYSCALL_H)) && \
+    (!(defined OS_MACOSX) && !(defined OS_OPENBSD))
 # define safe_write(fd, s, len)  syscall(SYS_write, fd, s, len)
 #else
  // Not so safe, but what can you do?
--- a/extern/glog/src/utilities.cc
+++ b/extern/glog/src/utilities.cc
@ -282,6 +282,8 @@ pid_t GetTID() {
  return getpid();  // Linux:  getpid returns thread ID when gettid is absent
 #elif defined OS_WINDOWS && !defined OS_CYGWIN
  return GetCurrentThreadId();
+#elif defined OS_OPENBSD
+  return getthrid();
 #else
  // If none of the techniques above worked, we use pthread_self().
  return (pid_t)(uintptr_t)pthread_self();
--- a/intern/CMakeLists.txt
+++ b/intern/CMakeLists.txt
@ -67,7 +67,7 @@ if(UNIX AND NOT APPLE)
  add_subdirectory(libc_compat)
 endif()

-if (WITH_RENDERDOC)
+if(WITH_RENDERDOC)
  add_subdirectory(renderdoc_dynload)
 endif()

--- a/intern/cycles/blender/shader.cpp
+++ b/intern/cycles/blender/shader.cpp
@ -523,27 +523,6 @@ static ShaderNode *add_node(Scene *scene,
  else if (b_node.is_a(&RNA_ShaderNodeHoldout)) {
    node = graph->create_node<HoldoutNode>();
  }
-  else if (b_node.is_a(&RNA_ShaderNodeBsdfAnisotropic)) {
-    BL::ShaderNodeBsdfAnisotropic b_aniso_node(b_node);
-    AnisotropicBsdfNode *aniso = graph->create_node<AnisotropicBsdfNode>();
-
-    switch (b_aniso_node.distribution()) {
-      case BL::ShaderNodeBsdfAnisotropic::distribution_BECKMANN:
-        aniso->set_distribution(CLOSURE_BSDF_MICROFACET_BECKMANN_ID);
-        break;
-      case BL::ShaderNodeBsdfAnisotropic::distribution_GGX:
-        aniso->set_distribution(CLOSURE_BSDF_MICROFACET_GGX_ID);
-        break;
-      case BL::ShaderNodeBsdfAnisotropic::distribution_MULTI_GGX:
-        aniso->set_distribution(CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID);
-        break;
-      case BL::ShaderNodeBsdfAnisotropic::distribution_ASHIKHMIN_SHIRLEY:
-        aniso->set_distribution(CLOSURE_BSDF_ASHIKHMIN_SHIRLEY_ID);
-        break;
-    }
-
-    node = aniso;
-  }
  else if (b_node.is_a(&RNA_ShaderNodeBsdfDiffuse)) {
    node = graph->create_node<DiffuseBsdfNode>();
  }
@ -566,24 +545,21 @@ static ShaderNode *add_node(Scene *scene,

    node = subsurface;
  }
-  else if (b_node.is_a(&RNA_ShaderNodeBsdfGlossy)) {
-    BL::ShaderNodeBsdfGlossy b_glossy_node(b_node);
+  else if (b_node.is_a(&RNA_ShaderNodeBsdfAnisotropic)) {
+    BL::ShaderNodeBsdfAnisotropic b_glossy_node(b_node);
    GlossyBsdfNode *glossy = graph->create_node<GlossyBsdfNode>();

    switch (b_glossy_node.distribution()) {
-      case BL::ShaderNodeBsdfGlossy::distribution_SHARP:
-        glossy->set_distribution(CLOSURE_BSDF_REFLECTION_ID);
-        break;
-      case BL::ShaderNodeBsdfGlossy::distribution_BECKMANN:
+      case BL::ShaderNodeBsdfAnisotropic::distribution_BECKMANN:
        glossy->set_distribution(CLOSURE_BSDF_MICROFACET_BECKMANN_ID);
        break;
-      case BL::ShaderNodeBsdfGlossy::distribution_GGX:
+      case BL::ShaderNodeBsdfAnisotropic::distribution_GGX:
        glossy->set_distribution(CLOSURE_BSDF_MICROFACET_GGX_ID);
        break;
-      case BL::ShaderNodeBsdfGlossy::distribution_ASHIKHMIN_SHIRLEY:
+      case BL::ShaderNodeBsdfAnisotropic::distribution_ASHIKHMIN_SHIRLEY:
        glossy->set_distribution(CLOSURE_BSDF_ASHIKHMIN_SHIRLEY_ID);
        break;
-      case BL::ShaderNodeBsdfGlossy::distribution_MULTI_GGX:
+      case BL::ShaderNodeBsdfAnisotropic::distribution_MULTI_GGX:
        glossy->set_distribution(CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID);
        break;
    }
@ -593,9 +569,6 @@ static ShaderNode *add_node(Scene *scene,
    BL::ShaderNodeBsdfGlass b_glass_node(b_node);
    GlassBsdfNode *glass = graph->create_node<GlassBsdfNode>();
    switch (b_glass_node.distribution()) {
-      case BL::ShaderNodeBsdfGlass::distribution_SHARP:
-        glass->set_distribution(CLOSURE_BSDF_SHARP_GLASS_ID);
-        break;
      case BL::ShaderNodeBsdfGlass::distribution_BECKMANN:
        glass->set_distribution(CLOSURE_BSDF_MICROFACET_BECKMANN_GLASS_ID);
        break;
@ -612,9 +585,6 @@ static ShaderNode *add_node(Scene *scene,
    BL::ShaderNodeBsdfRefraction b_refraction_node(b_node);
    RefractionBsdfNode *refraction = graph->create_node<RefractionBsdfNode>();
    switch (b_refraction_node.distribution()) {
-      case BL::ShaderNodeBsdfRefraction::distribution_SHARP:
-        refraction->set_distribution(CLOSURE_BSDF_REFRACTION_ID);
-        break;
      case BL::ShaderNodeBsdfRefraction::distribution_BECKMANN:
        refraction->set_distribution(CLOSURE_BSDF_MICROFACET_BECKMANN_REFRACTION_ID);
        break;
--- a/intern/cycles/device/oneapi/device_impl.cpp
+++ b/intern/cycles/device/oneapi/device_impl.cpp
@ -262,6 +262,11 @@ string OneapiDevice::oneapi_error_message()
  return string(oneapi_error_string_);
 }

+int OneapiDevice::scene_max_shaders()
+{
+  return scene_max_shaders_;
+}
+
 void *OneapiDevice::kernel_globals_device_pointer()
 {
  return kg_memory_device_;
@ -436,6 +441,9 @@ void OneapiDevice::const_copy_to(const char *name, void *host, size_t size)
    /* Update scene handle(since it is different for each device on multi devices) */
    KernelData *const data = (KernelData *)host;
    data->device_bvh = embree_scene;
+
+    /* We need this number later for proper local memory allocation. */
+    scene_max_shaders_ = data->max_shaders;
  }
 #  endif

--- a/intern/cycles/device/oneapi/device_impl.h
+++ b/intern/cycles/device/oneapi/device_impl.h
@ -37,6 +37,7 @@ class OneapiDevice : public Device {
  std::string oneapi_error_string_;
  bool use_hardware_raytracing = false;
  unsigned int kernel_features = 0;
+  int scene_max_shaders_ = 0;

 public:
  virtual BVHLayoutMask get_bvh_layout_mask(uint kernel_features) const override;
@ -61,6 +62,8 @@ class OneapiDevice : public Device {

  string oneapi_error_message();

+  int scene_max_shaders();
+
  void *kernel_globals_device_pointer();

  void mem_alloc(device_memory &mem) override;
--- a/intern/cycles/device/oneapi/queue.cpp
+++ b/intern/cycles/device/oneapi/queue.cpp
@ -59,7 +59,7 @@ void OneapiDeviceQueue::init_execution()
  void *kg_dptr = (void *)oneapi_device_->kernel_globals_device_pointer();
  assert(device_queue);
  assert(kg_dptr);
-  kernel_context_ = new KernelContext{device_queue, kg_dptr};
+  kernel_context_ = new KernelContext{device_queue, kg_dptr, 0};

  debug_init_execution();
 }
@ -78,12 +78,13 @@ bool OneapiDeviceQueue::enqueue(DeviceKernel kernel,
  assert(signed_kernel_work_size >= 0);
  size_t kernel_work_size = (size_t)signed_kernel_work_size;

+  assert(kernel_context_);
+  kernel_context_->scene_max_shaders = oneapi_device_->scene_max_shaders();
+
  size_t kernel_local_size = oneapi_kernel_preferred_local_size(
      kernel_context_->queue, (::DeviceKernel)kernel, kernel_work_size);
  size_t uniformed_kernel_work_size = round_up(kernel_work_size, kernel_local_size);

-  assert(kernel_context_);
-
  /* Call the oneAPI kernel DLL to launch the requested kernel. */
  bool is_finished_ok = oneapi_device_->enqueue_kernel(
      kernel_context_, kernel, uniformed_kernel_work_size, args);
--- a/intern/cycles/device/oneapi/queue.h
+++ b/intern/cycles/device/oneapi/queue.h
@ -39,6 +39,11 @@ class OneapiDeviceQueue : public DeviceQueue {
  virtual void copy_to_device(device_memory &mem) override;
  virtual void copy_from_device(device_memory &mem) override;

+  virtual bool supports_local_atomic_sort() const
+  {
+    return true;
+  }
+
 protected:
  OneapiDevice *oneapi_device_;
  KernelContext *kernel_context_;
--- a/intern/cycles/integrator/path_trace_work_gpu.cpp
+++ b/intern/cycles/integrator/path_trace_work_gpu.cpp
@ -386,11 +386,17 @@ void PathTraceWorkGPU::enqueue_reset()

  queue_->enqueue(DEVICE_KERNEL_INTEGRATOR_RESET, max_num_paths_, args);
  queue_->zero_to_device(integrator_queue_counter_);
-  queue_->zero_to_device(integrator_shader_sort_counter_);
-  if (device_scene_->data.kernel_features & KERNEL_FEATURE_NODE_RAYTRACE) {
+  if (integrator_shader_sort_counter_.size() != 0) {
+    queue_->zero_to_device(integrator_shader_sort_counter_);
+  }
+  if (device_scene_->data.kernel_features & KERNEL_FEATURE_NODE_RAYTRACE &&
+      integrator_shader_raytrace_sort_counter_.size() != 0)
+  {
    queue_->zero_to_device(integrator_shader_raytrace_sort_counter_);
  }
-  if (device_scene_->data.kernel_features & KERNEL_FEATURE_MNEE) {
+  if (device_scene_->data.kernel_features & KERNEL_FEATURE_MNEE &&
+      integrator_shader_mnee_sort_counter_.size() != 0)
+  {
    queue_->zero_to_device(integrator_shader_mnee_sort_counter_);
  }

--- a/intern/cycles/kernel/CMakeLists.txt
+++ b/intern/cycles/kernel/CMakeLists.txt
@ -847,6 +847,7 @@ if(WITH_CYCLES_DEVICE_ONEAPI)
    -DWITH_ONEAPI
    -ffast-math
    -O2
+    -D__KERNEL_LOCAL_ATOMIC_SORT__
    -o"${cycles_kernel_oneapi_lib}"
    -I"${CMAKE_CURRENT_SOURCE_DIR}/.."
    ${SYCL_CPP_FLAGS}
--- a/intern/cycles/kernel/device/gpu/kernel.h
+++ b/intern/cycles/kernel/device/gpu/kernel.h
@ -432,6 +432,17 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORTED_INDEX_DEFAULT_BLOCK_SIZE)
 }
 ccl_gpu_kernel_postfix

+/* oneAPI verion needs the local_mem accessor in the arguments. */
+#ifdef __KERNEL_ONEAPI__
+ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
+    ccl_gpu_kernel_signature(integrator_sort_bucket_pass,
+                             int num_states,
+                             int partition_size,
+                             int num_states_limit,
+                             ccl_global int *indices,
+                             int kernel_index,
+                             sycl::local_accessor<int> &local_mem)
+#else
 ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
    ccl_gpu_kernel_signature(integrator_sort_bucket_pass,
                             int num_states,
@ -439,9 +450,9 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
                             int num_states_limit,
                             ccl_global int *indices,
                             int kernel_index)
+#endif
 {
 #if defined(__KERNEL_LOCAL_ATOMIC_SORT__)
-  int max_shaders = context.launch_params_metal.data.max_shaders;
  ccl_global ushort *d_queued_kernel = (ccl_global ushort *)
                                           kernel_integrator_state.path.queued_kernel;
  ccl_global uint *d_shader_sort_key = (ccl_global uint *)
@ -449,6 +460,20 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
  ccl_global int *key_offsets = (ccl_global int *)
                                    kernel_integrator_state.sort_partition_key_offsets;

+#  ifdef __KERNEL_METAL__
+  int max_shaders = context.launch_params_metal.data.max_shaders;
+#  endif
+
+#  ifdef __KERNEL_ONEAPI__
+  /* Metal backend doesn't have these particular ccl_gpu_* defines and current kernel code
+   * uses metal_*, we need the below to be compatible with these kernels. */
+  int max_shaders = ((ONEAPIKernelContext *)kg)->__data->max_shaders;
+  int metal_local_id = ccl_gpu_thread_idx_x;
+  int metal_local_size = ccl_gpu_block_dim_x;
+  int metal_grid_id = ccl_gpu_block_idx_x;
+  ccl_gpu_shared int *threadgroup_array = local_mem.get_pointer();
+#  endif
+
  gpu_parallel_sort_bucket_pass(num_states,
                                partition_size,
                                max_shaders,
@ -456,7 +481,7 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
                                d_queued_kernel,
                                d_shader_sort_key,
                                key_offsets,
-                                (threadgroup int *)threadgroup_array,
+                                (ccl_gpu_shared int *)threadgroup_array,
                                metal_local_id,
                                metal_local_size,
                                metal_grid_id);
@ -464,6 +489,17 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
 }
 ccl_gpu_kernel_postfix

+/* oneAPI verion needs the local_mem accessor in the arguments. */
+#ifdef __KERNEL_ONEAPI__
+ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
+    ccl_gpu_kernel_signature(integrator_sort_write_pass,
+                             int num_states,
+                             int partition_size,
+                             int num_states_limit,
+                             ccl_global int *indices,
+                             int kernel_index,
+                             sycl::local_accessor<int> &local_mem)
+#else
 ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
    ccl_gpu_kernel_signature(integrator_sort_write_pass,
                             int num_states,
@ -471,9 +507,10 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
                             int num_states_limit,
                             ccl_global int *indices,
                             int kernel_index)
+#endif
+
 {
 #if defined(__KERNEL_LOCAL_ATOMIC_SORT__)
-  int max_shaders = context.launch_params_metal.data.max_shaders;
  ccl_global ushort *d_queued_kernel = (ccl_global ushort *)
                                           kernel_integrator_state.path.queued_kernel;
  ccl_global uint *d_shader_sort_key = (ccl_global uint *)
@ -481,6 +518,20 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
  ccl_global int *key_offsets = (ccl_global int *)
                                    kernel_integrator_state.sort_partition_key_offsets;

+#  ifdef __KERNEL_METAL__
+  int max_shaders = context.launch_params_metal.data.max_shaders;
+#  endif
+
+#  ifdef __KERNEL_ONEAPI__
+  /* Metal backend doesn't have these particular ccl_gpu_* defines and current kernel code
+   * uses metal_*, we need the below to be compatible with these kernels. */
+  int max_shaders = ((ONEAPIKernelContext *)kg)->__data->max_shaders;
+  int metal_local_id = ccl_gpu_thread_idx_x;
+  int metal_local_size = ccl_gpu_block_dim_x;
+  int metal_grid_id = ccl_gpu_block_idx_x;
+  ccl_gpu_shared int *threadgroup_array = local_mem.get_pointer();
+#  endif
+
  gpu_parallel_sort_write_pass(num_states,
                               partition_size,
                               max_shaders,
@ -490,7 +541,7 @@ ccl_gpu_kernel_threads(GPU_PARALLEL_SORT_BLOCK_SIZE)
                               d_queued_kernel,
                               d_shader_sort_key,
                               key_offsets,
-                               (threadgroup int *)threadgroup_array,
+                               (ccl_gpu_shared int *)threadgroup_array,
                               metal_local_id,
                               metal_local_size,
                               metal_grid_id);
--- a/intern/cycles/kernel/device/gpu/parallel_sorted_index.h
+++ b/intern/cycles/kernel/device/gpu/parallel_sorted_index.h
@ -23,11 +23,6 @@ CCL_NAMESPACE_BEGIN

 #if defined(__KERNEL_LOCAL_ATOMIC_SORT__)

-#  define atomic_store_local(p, x) \
-    atomic_store_explicit((threadgroup atomic_int *)p, x, memory_order_relaxed)
-#  define atomic_load_local(p) \
-    atomic_load_explicit((threadgroup atomic_int *)p, memory_order_relaxed)
-
 ccl_device_inline void gpu_parallel_sort_bucket_pass(const uint num_states,
                                                     const uint partition_size,
                                                     const uint max_shaders,
@ -45,7 +40,13 @@ ccl_device_inline void gpu_parallel_sort_bucket_pass(const uint num_states,
    atomic_store_local(&buckets[local_id], 0);
  }

+#  ifdef __KERNEL_ONEAPI__
+  /* NOTE(@nsirgien): For us here only local memory writing (buckets) is important,
+   * so faster local barriers can be used. */
+  ccl_gpu_local_syncthreads();
+#  else
  ccl_gpu_syncthreads();
+#  endif

  /* Determine bucket sizes within the partitions. */

@ -58,11 +59,17 @@ ccl_device_inline void gpu_parallel_sort_bucket_pass(const uint num_states,
    ushort kernel_index = d_queued_kernel[state_index];
    if (kernel_index == queued_kernel) {
      uint key = d_shader_sort_key[state_index] % max_shaders;
-      atomic_fetch_and_add_uint32(&buckets[key], 1);
+      atomic_fetch_and_add_uint32_shared(&buckets[key], 1);
    }
  }

+#  ifdef __KERNEL_ONEAPI__
+  /* NOTE(@nsirgien): For us here only local memory writing (buckets) is important,
+   * so faster local barriers can be used. */
+  ccl_gpu_local_syncthreads();
+#  else
  ccl_gpu_syncthreads();
+#  endif

  /* Calculate the partition's local offsets from the prefix sum of bucket sizes. */

@ -106,7 +113,13 @@ ccl_device_inline void gpu_parallel_sort_write_pass(const uint num_states,
    atomic_store_local(&local_offset[local_id], key_offsets[local_id] + partition_offset);
  }

+#  ifdef __KERNEL_ONEAPI__
+  /* NOTE(@nsirgien): For us here only local memory writing (local_offset) is important,
+   * so faster local barriers can be used. */
+  ccl_gpu_local_syncthreads();
+#  else
  ccl_gpu_syncthreads();
+#  endif

  /* Write the sorted active indices. */

@ -121,7 +134,7 @@ ccl_device_inline void gpu_parallel_sort_write_pass(const uint num_states,
    ushort kernel_index = d_queued_kernel[state_index];
    if (kernel_index == queued_kernel) {
      uint key = d_shader_sort_key[state_index] % max_shaders;
-      int index = atomic_fetch_and_add_uint32(&local_offset[key], 1);
+      int index = atomic_fetch_and_add_uint32_shared(&local_offset[key], 1);
      if (index < num_states_limit) {
        indices[index] = state_index;
      }
--- a/intern/cycles/kernel/device/oneapi/compat.h
+++ b/intern/cycles/kernel/device/oneapi/compat.h
@ -48,6 +48,7 @@
 #define ccl_loop_no_unroll
 #define ccl_optional_struct_init
 #define ccl_private
+#define ccl_gpu_shared
 #define ATTR_FALLTHROUGH __attribute__((fallthrough))
 #define ccl_constant const
 #define ccl_try_align(...) __attribute__((aligned(__VA_ARGS__)))
--- a/intern/cycles/kernel/device/oneapi/context_begin.h
+++ b/intern/cycles/kernel/device/oneapi/context_begin.h
@ -2,8 +2,40 @@
 * Copyright 2021-2022 Intel Corporation */

 #ifdef WITH_NANOVDB
+/* Data type to replace `double` used in the NanoVDB headers. Cycles don't need doubles, and is
+ * safer and more portable to never use double datatype on GPU.
+ * Use a special structure, so that the following is true:
+ * - No unnoticed implicit cast or mathematical operations used on scalar 64bit type
+ *   (which rules out trick like using `uint64_t` as a drop-in replacement for double).
+ * - Padding rules are matching exactly `double`
+ *   (which rules out array of `uint8_t`). */
+typedef struct ccl_vdb_double_t {
+  union ccl_vdb_helper_t {
+    double d;
+    uint64_t i;
+  };
+
+  uint64_t i;
+  ccl_vdb_double_t(double value)
+  {
+    ccl_vdb_helper_t helper;
+    helper.d = value;
+    i = helper.i;
+  }
+  /* We intentionally allow conversion to float in order to workaround compilation errors
+   * for defined math functions that take doubles. */
+  operator float() const
+  {
+    ccl_vdb_helper_t helper;
+    helper.i = i;
+    return (float)helper.d;
+  }
+} ccl_vdb_double_t;
+
+#  define double ccl_vdb_double_t
 #  include <nanovdb/NanoVDB.h>
 #  include <nanovdb/util/SampleFromVoxels.h>
+#  undef double
 #endif

 /* clang-format off */
--- a/intern/cycles/kernel/device/oneapi/kernel.cpp
+++ b/intern/cycles/kernel/device/oneapi/kernel.cpp
@ -109,7 +109,10 @@ size_t oneapi_kernel_preferred_local_size(SyclQueue *queue,
  assert(queue);
  (void)kernel_global_size;
  const static size_t preferred_work_group_size_intersect_shading = 32;
-  const static size_t preferred_work_group_size_technical = 1024;
+  /* Shader evalutation kernels seems to use some amount of shared memory, so better
+   * to avoid usage of maximum work group sizes for them. */
+  const static size_t preferred_work_group_size_shader_evaluation = 256;
+  const static size_t preferred_work_group_size_default = 1024;

  size_t preferred_work_group_size = 0;
  switch (kernel) {
@ -133,19 +136,36 @@ size_t oneapi_kernel_preferred_local_size(SyclQueue *queue,
    case DEVICE_KERNEL_INTEGRATOR_QUEUED_SHADOW_PATHS_ARRAY:
    case DEVICE_KERNEL_INTEGRATOR_ACTIVE_PATHS_ARRAY:
    case DEVICE_KERNEL_INTEGRATOR_TERMINATED_PATHS_ARRAY:
-    case DEVICE_KERNEL_INTEGRATOR_SORTED_PATHS_ARRAY:
-    case DEVICE_KERNEL_INTEGRATOR_COMPACT_PATHS_ARRAY:
-    case DEVICE_KERNEL_INTEGRATOR_COMPACT_STATES:
    case DEVICE_KERNEL_INTEGRATOR_TERMINATED_SHADOW_PATHS_ARRAY:
+    case DEVICE_KERNEL_INTEGRATOR_COMPACT_PATHS_ARRAY:
    case DEVICE_KERNEL_INTEGRATOR_COMPACT_SHADOW_PATHS_ARRAY:
+      preferred_work_group_size = GPU_PARALLEL_ACTIVE_INDEX_DEFAULT_BLOCK_SIZE;
+      break;
+
+    case DEVICE_KERNEL_INTEGRATOR_SORTED_PATHS_ARRAY:
+    case DEVICE_KERNEL_INTEGRATOR_COMPACT_STATES:
    case DEVICE_KERNEL_INTEGRATOR_COMPACT_SHADOW_STATES:
-    case DEVICE_KERNEL_INTEGRATOR_RESET:
-    case DEVICE_KERNEL_INTEGRATOR_SHADOW_CATCHER_COUNT_POSSIBLE_SPLITS:
-      preferred_work_group_size = preferred_work_group_size_technical;
+      preferred_work_group_size = GPU_PARALLEL_SORTED_INDEX_DEFAULT_BLOCK_SIZE;
+      break;
+
+    case DEVICE_KERNEL_INTEGRATOR_SORT_BUCKET_PASS:
+    case DEVICE_KERNEL_INTEGRATOR_SORT_WRITE_PASS:
+      preferred_work_group_size = GPU_PARALLEL_SORT_BLOCK_SIZE;
+      break;
+
+    case DEVICE_KERNEL_PREFIX_SUM:
+      preferred_work_group_size = GPU_PARALLEL_PREFIX_SUM_DEFAULT_BLOCK_SIZE;
+      break;
+
+    case DEVICE_KERNEL_SHADER_EVAL_DISPLACE:
+    case DEVICE_KERNEL_SHADER_EVAL_BACKGROUND:
+    case DEVICE_KERNEL_SHADER_EVAL_CURVE_SHADOW_TRANSPARENCY:
+      preferred_work_group_size = preferred_work_group_size_shader_evaluation;
      break;

    default:
-      preferred_work_group_size = 512;
+      preferred_work_group_size = preferred_work_group_size_default;
+      break;
  }

  const size_t limit_work_group_size = reinterpret_cast<sycl::queue *>(queue)
@ -316,12 +336,6 @@ bool oneapi_enqueue_kernel(KernelContext *kernel_context,
      kernel_context->queue, device_kernel, global_size);
  assert(global_size % local_size == 0);

-  /* Local size for DEVICE_KERNEL_INTEGRATOR_ACTIVE_PATHS_ARRAY needs to be enforced so we
-   * overwrite it outside of oneapi_kernel_preferred_local_size. */
-  if (device_kernel == DEVICE_KERNEL_INTEGRATOR_ACTIVE_PATHS_ARRAY) {
-    local_size = GPU_PARALLEL_ACTIVE_INDEX_DEFAULT_BLOCK_SIZE;
-  }
-
  /* Kernels listed below need a specific number of work groups. */
  if (device_kernel == DEVICE_KERNEL_INTEGRATOR_ACTIVE_PATHS_ARRAY ||
      device_kernel == DEVICE_KERNEL_INTEGRATOR_QUEUED_PATHS_ARRAY ||
@ -353,6 +367,14 @@ bool oneapi_enqueue_kernel(KernelContext *kernel_context,
 #    pragma GCC diagnostic error "-Wswitch"
 #  endif

+  int max_shaders = 0;
+
+  if (device_kernel == DEVICE_KERNEL_INTEGRATOR_SORT_BUCKET_PASS ||
+      device_kernel == DEVICE_KERNEL_INTEGRATOR_SORT_WRITE_PASS)
+  {
+    max_shaders = (kernel_context->scene_max_shaders);
+  }
+
  try {
    queue->submit([&](sycl::handler &cgh) {
 #  ifdef WITH_EMBREE_GPU
@ -495,13 +517,31 @@ bool oneapi_enqueue_kernel(KernelContext *kernel_context,
          break;
        }
        case DEVICE_KERNEL_INTEGRATOR_SORT_BUCKET_PASS: {
-          oneapi_call(
-              kg, cgh, global_size, local_size, args, oneapi_kernel_integrator_sort_bucket_pass);
+          sycl::local_accessor<int> local_mem(max_shaders, cgh);
+          oneapi_kernel_integrator_sort_bucket_pass(kg,
+                                                    global_size,
+                                                    local_size,
+                                                    cgh,
+                                                    *(int *)(args[0]),
+                                                    *(int *)(args[1]),
+                                                    *(int *)(args[2]),
+                                                    *(int **)(args[3]),
+                                                    *(int *)(args[4]),
+                                                    local_mem);
          break;
        }
        case DEVICE_KERNEL_INTEGRATOR_SORT_WRITE_PASS: {
-          oneapi_call(
-              kg, cgh, global_size, local_size, args, oneapi_kernel_integrator_sort_write_pass);
+          sycl::local_accessor<int> local_mem(max_shaders, cgh);
+          oneapi_kernel_integrator_sort_write_pass(kg,
+                                                   global_size,
+                                                   local_size,
+                                                   cgh,
+                                                   *(int *)(args[0]),
+                                                   *(int *)(args[1]),
+                                                   *(int *)(args[2]),
+                                                   *(int **)(args[3]),
+                                                   *(int *)(args[4]),
+                                                   local_mem);
          break;
        }
        case DEVICE_KERNEL_INTEGRATOR_COMPACT_PATHS_ARRAY: {
--- a/intern/cycles/kernel/device/oneapi/kernel.h
+++ b/intern/cycles/kernel/device/oneapi/kernel.h
@ -32,6 +32,8 @@ struct KernelContext {
  SyclQueue *queue;
  /* Pointer to USM device memory with all global/constant allocation on this device */
  void *kernel_globals;
+  /* We needs this additional data for some kernels. */
+  int scene_max_shaders;
 };

 /* Use extern C linking so that the symbols can be easily load from the dynamic library at runtime.
--- a/intern/cycles/kernel/osl/shaders/CMakeLists.txt
+++ b/intern/cycles/kernel/osl/shaders/CMakeLists.txt
@ -6,7 +6,6 @@
 set(SRC_OSL
  node_add_closure.osl
  node_ambient_occlusion.osl
-  node_anisotropic_bsdf.osl
  node_attribute.osl
  node_background.osl
  node_bevel.osl
--- a/intern/cycles/kernel/osl/shaders/node_anisotropic_bsdf.osl
+++ b/intern/cycles/kernel/osl/shaders/node_anisotropic_bsdf.osl
@ -1,45 +0,0 @@
-/* SPDX-License-Identifier: Apache-2.0
- * Copyright 2011-2022 Blender Foundation */
-
-#include "stdcycles.h"
-
-shader node_anisotropic_bsdf(color Color = 0.0,
-                             string distribution = "GGX",
-                             float Roughness = 0.0,
-                             float Anisotropy = 0.0,
-                             float Rotation = 0.0,
-                             normal Normal = N,
-                             normal Tangent = normalize(dPdu),
-                             output closure color BSDF = 0)
-{
-  /* rotate tangent around normal */
-  vector T = Tangent;
-
-  if (Rotation != 0.0)
-    T = rotate(T, Rotation * M_2PI, point(0.0, 0.0, 0.0), Normal);
-
-  /* compute roughness */
-  float roughness = Roughness * Roughness;
-  float roughness_u, roughness_v;
-  float aniso = clamp(Anisotropy, -0.99, 0.99);
-
-  if (aniso < 0.0) {
-    roughness_u = roughness / (1.0 + aniso);
-    roughness_v = roughness * (1.0 + aniso);
-  }
-  else {
-    roughness_u = roughness * (1.0 - aniso);
-    roughness_v = roughness / (1.0 - aniso);
-  }
-
-  if (distribution == "sharp")
-    BSDF = Color * reflection(Normal);
-  else if (distribution == "beckmann")
-    BSDF = Color * microfacet_beckmann_aniso(Normal, T, roughness_u, roughness_v);
-  else if (distribution == "GGX")
-    BSDF = Color * microfacet_ggx_aniso(Normal, T, roughness_u, roughness_v);
-  else if (distribution == "Multiscatter GGX")
-    BSDF = Color * microfacet_multi_ggx_aniso(Normal, T, roughness_u, roughness_v, Color);
-  else
-    BSDF = Color * ashikhmin_shirley(Normal, T, roughness_u, roughness_v);
-}
--- a/intern/cycles/kernel/osl/shaders/node_glossy_bsdf.osl
+++ b/intern/cycles/kernel/osl/shaders/node_glossy_bsdf.osl
@ -7,13 +7,40 @@
 shader node_glossy_bsdf(color Color = 0.8,
                        string distribution = "ggx",
                        float Roughness = 0.2,
+                        float Anisotropy = 0.0,
+                        float Rotation = 0.0,
                        normal Normal = N,
+                        normal Tangent = 0.0,
                        output closure color BSDF = 0)
 {
+  /* compute roughness */
  float roughness = Roughness * Roughness;
+  float roughness_u, roughness_v;
+  float aniso = clamp(Anisotropy, -0.99, 0.99);
+
+  /* rotate tangent around normal */
+  vector T = Tangent;
+
+  if (abs(aniso) <= 1e-4) {
+    roughness_u = roughness;
+    roughness_v = roughness;
+  }
+  else {
+    if (Rotation != 0.0)
+      T = rotate(T, Rotation * M_2PI, point(0.0, 0.0, 0.0), Normal);
+
+    if (aniso < 0.0) {
+      roughness_u = roughness / (1.0 + aniso);
+      roughness_v = roughness * (1.0 + aniso);
+    }
+    else {
+      roughness_u = roughness * (1.0 - aniso);
+      roughness_v = roughness / (1.0 - aniso);
+    }
+  }

  if (distribution == "Multiscatter GGX")
    BSDF = Color * microfacet_multi_ggx(Normal, roughness, Color);
  else
-    BSDF = Color * microfacet(distribution, Normal, roughness, 0.0, 0);
+    BSDF = Color * microfacet(distribution, Normal, T, roughness_u, roughness_v, 0.0, 0);
 }
--- a/intern/cycles/kernel/svm/closure.h
+++ b/intern/cycles/kernel/svm/closure.h
@ -494,7 +494,10 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
      bsdf->ior = 1.0f;
      bsdf->fresnel = NULL;

-      if (data_node.y == SVM_STACK_INVALID) {
+      /* compute roughness */
+      float anisotropy = clamp(param2, -0.99f, 0.99f);
+      if (data_node.y == SVM_STACK_INVALID || fabsf(anisotropy) <= 1e-4f) {
+        /* Isotropic case. */
        bsdf->T = zero_float3();
        bsdf->alpha_x = roughness;
        bsdf->alpha_y = roughness;
@ -507,8 +510,6 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
        if (rotation != 0.0f)
          bsdf->T = rotate_around_axis(bsdf->T, bsdf->N, rotation * M_2PI_F);

-        /* compute roughness */
-        float anisotropy = clamp(param2, -0.99f, 0.99f);
        if (anisotropy < 0.0f) {
          bsdf->alpha_x = roughness / (1.0f + anisotropy);
          bsdf->alpha_y = roughness * (1.0f + anisotropy);
--- a/intern/cycles/scene/shader_nodes.cpp
+++ b/intern/cycles/scene/shader_nodes.cpp
@ -2335,68 +2335,6 @@ void BsdfNode::compile(OSLCompiler & /*compiler*/)
  assert(0);
 }

-/* Anisotropic BSDF Closure */
-
-NODE_DEFINE(AnisotropicBsdfNode)
-{
-  NodeType *type = NodeType::add("anisotropic_bsdf", create, NodeType::SHADER);
-
-  SOCKET_IN_COLOR(color, "Color", make_float3(0.8f, 0.8f, 0.8f));
-  SOCKET_IN_NORMAL(normal, "Normal", zero_float3(), SocketType::LINK_NORMAL);
-  SOCKET_IN_FLOAT(surface_mix_weight, "SurfaceMixWeight", 0.0f, SocketType::SVM_INTERNAL);
-
-  static NodeEnum distribution_enum;
-  distribution_enum.insert("beckmann", CLOSURE_BSDF_MICROFACET_BECKMANN_ID);
-  distribution_enum.insert("GGX", CLOSURE_BSDF_MICROFACET_GGX_ID);
-  distribution_enum.insert("Multiscatter GGX", CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID);
-  distribution_enum.insert("ashikhmin_shirley", CLOSURE_BSDF_ASHIKHMIN_SHIRLEY_ID);
-  SOCKET_ENUM(distribution, "Distribution", distribution_enum, CLOSURE_BSDF_MICROFACET_GGX_ID);
-
-  SOCKET_IN_VECTOR(tangent, "Tangent", zero_float3(), SocketType::LINK_TANGENT);
-
-  SOCKET_IN_FLOAT(roughness, "Roughness", 0.5f);
-  SOCKET_IN_FLOAT(anisotropy, "Anisotropy", 0.5f);
-  SOCKET_IN_FLOAT(rotation, "Rotation", 0.0f);
-
-  SOCKET_OUT_CLOSURE(BSDF, "BSDF");
-
-  return type;
-}
-
-AnisotropicBsdfNode::AnisotropicBsdfNode() : BsdfNode(get_node_type())
-{
-  closure = CLOSURE_BSDF_MICROFACET_GGX_ID;
-}
-
-void AnisotropicBsdfNode::attributes(Shader *shader, AttributeRequestSet *attributes)
-{
-  if (shader->has_surface_link()) {
-    ShaderInput *tangent_in = input("Tangent");
-
-    if (!tangent_in->link)
-      attributes->add(ATTR_STD_GENERATED);
-  }
-
-  ShaderNode::attributes(shader, attributes);
-}
-
-void AnisotropicBsdfNode::compile(SVMCompiler &compiler)
-{
-  closure = distribution;
-
-  if (closure == CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID)
-    BsdfNode::compile(
-        compiler, input("Roughness"), input("Anisotropy"), input("Rotation"), input("Color"));
-  else
-    BsdfNode::compile(compiler, input("Roughness"), input("Anisotropy"), input("Rotation"));
-}
-
-void AnisotropicBsdfNode::compile(OSLCompiler &compiler)
-{
-  compiler.parameter(this, "distribution");
-  compiler.add(this, "node_anisotropic_bsdf");
-}
-
 /* Glossy BSDF Closure */

 NODE_DEFINE(GlossyBsdfNode)
@ -2414,7 +2352,12 @@ NODE_DEFINE(GlossyBsdfNode)
  distribution_enum.insert("ashikhmin_shirley", CLOSURE_BSDF_ASHIKHMIN_SHIRLEY_ID);
  distribution_enum.insert("Multiscatter GGX", CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID);
  SOCKET_ENUM(distribution, "Distribution", distribution_enum, CLOSURE_BSDF_MICROFACET_GGX_ID);
+
+  SOCKET_IN_VECTOR(tangent, "Tangent", zero_float3(), SocketType::LINK_TANGENT);
+
  SOCKET_IN_FLOAT(roughness, "Roughness", 0.5f);
+  SOCKET_IN_FLOAT(anisotropy, "Anisotropy", 0.0f);
+  SOCKET_IN_FLOAT(rotation, "Rotation", 0.0f);

  SOCKET_OUT_CLOSURE(BSDF, "BSDF");

@ -2427,8 +2370,32 @@ GlossyBsdfNode::GlossyBsdfNode() : BsdfNode(get_node_type())
  distribution_orig = NBUILTIN_CLOSURES;
 }

+bool GlossyBsdfNode::is_isotropic()
+{
+  ShaderInput *anisotropy_input = input("Anisotropy");
+  /* Keep in sync with the thresholds in OSL's node_glossy_bsdf and SVM's svm_node_closure_bsdf. */
+  return (!anisotropy_input->link && fabsf(anisotropy) <= 1e-4f);
+}
+
+void GlossyBsdfNode::attributes(Shader *shader, AttributeRequestSet *attributes)
+{
+  if (shader->has_surface_link()) {
+    ShaderInput *tangent_in = input("Tangent");
+    if (!tangent_in->link && !is_isotropic())
+      attributes->add(ATTR_STD_GENERATED);
+  }
+
+  ShaderNode::attributes(shader, attributes);
+}
+
 void GlossyBsdfNode::simplify_settings(Scene *scene)
 {
+  /* If the anisotropy is close enough to zero, fall back to the isotropic case. */
+  ShaderInput *tangent_input = input("Tangent");
+  if (tangent_input->link && is_isotropic()) {
+    tangent_input->disconnect();
+  }
+
  if (distribution_orig == NBUILTIN_CLOSURES) {
    roughness_orig = roughness;
    distribution_orig = distribution;
@ -2478,9 +2445,10 @@ void GlossyBsdfNode::compile(SVMCompiler &compiler)
  if (closure == CLOSURE_BSDF_REFLECTION_ID)
    BsdfNode::compile(compiler, NULL, NULL);
  else if (closure == CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID)
-    BsdfNode::compile(compiler, input("Roughness"), NULL, NULL, input("Color"));
+    BsdfNode::compile(
+        compiler, input("Roughness"), input("Anisotropy"), input("Rotation"), input("Color"));
  else
-    BsdfNode::compile(compiler, input("Roughness"), NULL);
+    BsdfNode::compile(compiler, input("Roughness"), input("Anisotropy"), input("Rotation"));
 }

 void GlossyBsdfNode::compile(OSLCompiler &compiler)
--- a/intern/cycles/scene/shader_nodes.h
+++ b/intern/cycles/scene/shader_nodes.h
@ -495,27 +495,6 @@ class BsdfNode : public BsdfBaseNode {
  NODE_SOCKET_API(float, surface_mix_weight)
 };

-class AnisotropicBsdfNode : public BsdfNode {
- public:
-  SHADER_NODE_CLASS(AnisotropicBsdfNode)
-
-  NODE_SOCKET_API(float3, tangent)
-  NODE_SOCKET_API(float, roughness)
-  NODE_SOCKET_API(float, anisotropy)
-  NODE_SOCKET_API(float, rotation)
-  NODE_SOCKET_API(ClosureType, distribution)
-
-  ClosureType get_closure_type()
-  {
-    return distribution;
-  }
-  void attributes(Shader *shader, AttributeRequestSet *attributes);
-  bool has_attribute_dependency()
-  {
-    return true;
-  }
-};
-
 class DiffuseBsdfNode : public BsdfNode {
 public:
  SHADER_NODE_CLASS(DiffuseBsdfNode)
@ -624,12 +603,23 @@ class GlossyBsdfNode : public BsdfNode {
    return distribution;
  }

+  NODE_SOCKET_API(float3, tangent)
  NODE_SOCKET_API(float, roughness)
+  NODE_SOCKET_API(float, anisotropy)
+  NODE_SOCKET_API(float, rotation)
  NODE_SOCKET_API(ClosureType, distribution)

+  void attributes(Shader *shader, AttributeRequestSet *attributes);
+  bool has_attribute_dependency()
+  {
+    return true;
+  }
+
 private:
  float roughness_orig;
  ClosureType distribution_orig;
+
+  bool is_isotropic();
 };

 class GlassBsdfNode : public BsdfNode {
--- a/intern/cycles/util/atomic.h
+++ b/intern/cycles/util/atomic.h
@ -21,6 +21,10 @@

 #else /* __KERNEL_GPU__ */

+#  ifndef __KERNEL_ONEAPI__
+#    define atomic_fetch_and_add_uint32_shared atomic_fetch_and_add_uint32
+#  endif
+
 #  if defined(__KERNEL_CUDA__) || defined(__KERNEL_HIP__)

 #    define atomic_add_and_fetch_float(p, x) (atomicAdd((float *)(p), (float)(x)) + (float)(x))
@ -140,6 +144,11 @@ ccl_device_inline float atomic_compare_and_swap_float(volatile ccl_global float
 #    define atomic_store(p, x) atomic_store_explicit(p, x, memory_order_relaxed)
 #    define atomic_fetch(p) atomic_load_explicit(p, memory_order_relaxed)

+#    define atomic_store_local(p, x) \
+      atomic_store_explicit((ccl_gpu_shared atomic_int *)p, x, memory_order_relaxed)
+#    define atomic_load_local(p) \
+      atomic_load_explicit((ccl_gpu_shared atomic_int *)p, memory_order_relaxed)
+
 #    define CCL_LOCAL_MEM_FENCE mem_flags::mem_threadgroup
 #    define ccl_barrier(flags) threadgroup_barrier(flags)

@ -191,6 +200,16 @@ ccl_device_inline int atomic_fetch_and_add_uint32(ccl_global int *p, int x)
  return atomic.fetch_add(x);
 }

+ccl_device_inline int atomic_fetch_and_add_uint32_shared(int *p, int x)
+{
+  sycl::atomic_ref<int,
+                   sycl::memory_order::relaxed,
+                   sycl::memory_scope::device,
+                   sycl::access::address_space::local_space>
+      atomic(*p);
+  return atomic.fetch_add(x);
+}
+
 ccl_device_inline unsigned int atomic_fetch_and_sub_uint32(ccl_global unsigned int *p,
                                                           unsigned int x)
 {
@ -253,6 +272,26 @@ ccl_device_inline int atomic_fetch_and_or_uint32(ccl_global int *p, int x)
  return atomic.fetch_or(x);
 }

+ccl_device_inline void atomic_store_local(int *p, int x)
+{
+  sycl::atomic_ref<int,
+                   sycl::memory_order::relaxed,
+                   sycl::memory_scope::device,
+                   sycl::access::address_space::local_space>
+      atomic(*p);
+  atomic.store(x);
+}
+
+ccl_device_inline int atomic_load_local(int *p)
+{
+  sycl::atomic_ref<int,
+                   sycl::memory_order::relaxed,
+                   sycl::memory_scope::device,
+                   sycl::access::address_space::local_space>
+      atomic(*p);
+  return atomic.load();
+}
+
 #  endif /* __KERNEL_ONEAPI__ */

 #endif /* __KERNEL_GPU__ */
--- a/intern/cycles/util/math_float2.h
+++ b/intern/cycles/util/math_float2.h
@ -134,6 +134,21 @@ ccl_device_inline float len(const float2 a)
  return sqrtf(dot(a, a));
 }

+ccl_device_inline float reduce_min(const float2 a)
+{
+  return min(a.x, a.y);
+}
+
+ccl_device_inline float reduce_max(const float2 a)
+{
+  return max(a.x, a.y);
+}
+
+ccl_device_inline float reduce_add(const float2 a)
+{
+  return a.x + a.y;
+}
+
 ccl_device_inline float len_squared(const float2 a)
 {
  return dot(a, a);
--- a/intern/ghost/intern/GHOST_SystemCocoa.mm
+++ b/intern/ghost/intern/GHOST_SystemCocoa.mm
@ -1261,7 +1261,7 @@ GHOST_TSuccess GHOST_SystemCocoa::handleDraggingEvent(GHOST_TEventType eventType
              ![bitmapImage isPlanar])
          {
            /* Try a fast copy if the image is a meshed RGBA 32bit bitmap. */
-            toIBuf = (uint8_t *)ibuf->rect;
+            toIBuf = ibuf->byte_buffer.data;
            rasterRGB = (uint8_t *)[bitmapImage bitmapData];
            for (y = 0; y < imgSize.height; y++) {
              to_i = (imgSize.height - y - 1) * imgSize.width;
@ -1338,7 +1338,7 @@ GHOST_TSuccess GHOST_SystemCocoa::handleDraggingEvent(GHOST_TEventType eventType
            }

            /* Copy the image to ibuf, flipping it vertically. */
-            toIBuf = (uint8_t *)ibuf->rect;
+            toIBuf = ibuf->byte_buffer.data;
            for (y = 0; y < imgSize.height; y++) {
              for (x = 0; x < imgSize.width; x++) {
                to_i = (imgSize.height - y - 1) * imgSize.width + x;
--- a/intern/ghost/intern/GHOST_SystemWayland.cc
+++ b/intern/ghost/intern/GHOST_SystemWayland.cc
@ -173,7 +173,7 @@ static bool use_gnome_confine_hack = false;
 * This is a hack because it seems there is no way to check if the compositor supports
 * server side decorations when initializing WAYLAND.
 */
-#if defined(WITH_GHOST_WAYLAND_LIBDECOR) && defined(WITH_GHOST_X11)
+#ifdef WITH_GHOST_WAYLAND_LIBDECOR
 #  define USE_GNOME_NEEDS_LIBDECOR_HACK
 #endif

@ -1333,6 +1333,22 @@ static void ghost_wayland_log_handler(const char *msg, va_list arg)
  }
 }

+#ifdef WITH_GHOST_X11
+/**
+ * Check if the system is running X11.
+ * This is not intended to be a fool-proof check (the `DISPLAY` is not validated for e.g.).
+ * Just check `DISPLAY` is set and not-empty.
+ */
+static bool ghost_wayland_is_x11_available()
+{
+  const char *x11_display = getenv("DISPLAY");
+  if (x11_display && x11_display[0]) {
+    return true;
+  }
+  return false;
+}
+#endif /* WITH_GHOST_X11 */
+
 static GHOST_TKey xkb_map_gkey(const xkb_keysym_t sym)
 {

@ -5517,12 +5533,21 @@ GHOST_SystemWayland::GHOST_SystemWayland(bool background)
  }

 #ifdef WITH_GHOST_WAYLAND_LIBDECOR
-  /* Ignore windowing requirements when running in background mode,
-   * as it doesn't make sense to fall back to X11 because of windowing functionality
-   * in background mode, also LIBDECOR is crashing in background mode `blender -b -f 1`
-   * for e.g. while it could be fixed, requiring the library at all makes no sense . */
-  if (background) {
-    display_->libdecor_required = false;
+  if (display_->libdecor_required) {
+    /* Ignore windowing requirements when running in background mode,
+     * as it doesn't make sense to fall back to X11 because of windowing functionality
+     * in background mode, also LIBDECOR is crashing in background mode `blender -b -f 1`
+     * for e.g. while it could be fixed, requiring the library at all makes no sense . */
+    if (background) {
+      display_->libdecor_required = false;
+    }
+#  ifdef WITH_GHOST_X11
+    else if (!has_libdecor && !ghost_wayland_is_x11_available()) {
+      /* Only require LIBDECOR when X11 is available, otherwise there is nothing to fall back to.
+       * It's better to open without window decorations than failing entirely. */
+      display_->libdecor_required = false;
+    }
+#  endif /* WITH_GHOST_X11 */
  }

  if (display_->libdecor_required) {
--- a/intern/ghost/intern/GHOST_SystemWin32.cc
+++ b/intern/ghost/intern/GHOST_SystemWin32.cc
@ -2422,7 +2422,7 @@ static uint *getClipboardImageImBuf(int *r_width, int *r_height, UINT format)
    *r_width = ibuf->x;
    *r_height = ibuf->y;
    rgba = (uint *)malloc(4 * ibuf->x * ibuf->y);
-    memcpy(rgba, ibuf->rect, 4 * ibuf->x * ibuf->y);
+    memcpy(rgba, ibuf->byte_buffer.data, 4 * ibuf->x * ibuf->y);
    IMB_freeImBuf(ibuf);
  }

@ -2513,7 +2513,7 @@ static bool putClipboardImagePNG(uint *rgba, int width, int height)
  UINT cf = RegisterClipboardFormat("PNG");

  /* Load buffer into ImBuf, convert to PNG. */
-  ImBuf *ibuf = IMB_allocFromBuffer(rgba, nullptr, width, height, 32);
+  ImBuf *ibuf = IMB_allocFromBuffer(reinterpret_cast<uint8_t *>(rgba), nullptr, width, height, 32);
  ibuf->ftype = IMB_FTYPE_PNG;
  ibuf->foptions.quality = 15;
  if (!IMB_saveiff(ibuf, "<memory>", IB_rect | IB_mem)) {
@ -2521,7 +2521,7 @@ static bool putClipboardImagePNG(uint *rgba, int width, int height)
    return false;
  }

-  HGLOBAL hMem = GlobalAlloc(GHND, ibuf->encodedbuffersize);
+  HGLOBAL hMem = GlobalAlloc(GHND, ibuf->encoded_buffer_size);
  if (!hMem) {
    IMB_freeImBuf(ibuf);
    return false;
@ -2534,7 +2534,7 @@ static bool putClipboardImagePNG(uint *rgba, int width, int height)
    return false;
  }

-  memcpy(pMem, ibuf->encodedbuffer, ibuf->encodedbuffersize);
+  memcpy(pMem, ibuf->encoded_buffer.data, ibuf->encoded_buffer_size);

  GlobalUnlock(hMem);
  IMB_freeImBuf(ibuf);
--- a/intern/renderdoc_dynload/CMakeLists.txt
+++ b/intern/renderdoc_dynload/CMakeLists.txt
@ -9,9 +9,9 @@ set(INC_SYS
 )

 set(SRC
-    intern/renderdoc_api.cc
+  intern/renderdoc_api.cc

-    include/renderdoc_api.hh
+  include/renderdoc_api.hh
 )

 blender_add_lib(bf_intern_renderdoc_dynload "${SRC}" "${INC}" "${INC_SYS}" "${LIB}")
--- a/release/datafiles/startup.blend
+++ b/release/datafiles/startup.blend
--- a/scripts/modules/bl_i18n_utils/bl_extract_messages.py
+++ b/scripts/modules/bl_i18n_utils/bl_extract_messages.py
@ -204,7 +204,7 @@ def dump_rna_messages(msgs, reports, settings, verbose=False):
    def class_blacklist():
        blacklist_rna_class = {getattr(bpy.types, cls_id) for cls_id in (
            # core classes
-            "Context", "Event", "Function", "UILayout", "UnknownType", "Property", "Struct",
+            "Context", "Event", "Function", "UILayout", "UnknownType", "Struct",
            # registerable classes
            "Panel", "Menu", "Header", "RenderEngine", "Operator", "OperatorMacro", "Macro", "KeyingSetInfo",
        )
--- a/scripts/modules/bl_i18n_utils/settings.py
+++ b/scripts/modules/bl_i18n_utils/settings.py
@ -256,6 +256,25 @@ PYGETTEXT_KEYWORDS = (() +
    tuple(("{}\\((?:[^,]+,){{2}}\\s*" + _msg_re + r"\s*(?:\)|,)").format(it)
          for it in ("modifier_subpanel_register", "gpencil_modifier_subpanel_register")) +

+    # Node socket declarations: contextless names
+    tuple((r"\.{}<decl::.*?>\(\s*" + _msg_re + r"(?:,[^),]+)*\s*\)").format(it)
+          for it in ("add_input", "add_output")) +
+
+    # Node socket declarations: names with contexts
+    tuple((r"\.{}<decl::.*?>\(\s*" + _msg_re + r"[^;]*\.translation_context\(\s*" + _ctxt_re + r"\s*\)").format(it)
+          for it in ("add_input", "add_output")) +
+
+    # Node socket declarations: description and error messages
+    tuple((r"\.{}\(\s*" + _msg_re + r"\s*\)").format(it)
+          for it in ("description", "error_message_add")) +
+
+    # Node socket labels
+    tuple((r"{}\(\s*[^,]+,\s*" + _msg_re + r"\s*\)").format(it)
+          for it in ("node_sock_label",)) +
+
+    # Geometry Nodes field inputs
+    ((r"FieldInput\(CPPType::get<.*?>\(\),\s*" + _msg_re + r"\s*\)"),) +
+
    # bUnitDef unit names.
    # NOTE: regex is a bit more complex than it would need too. Since the actual
    # identifier (`B_UNIT_DEF_`) is at the end, if it's simpler/too general it
--- a/scripts/presets/keyconfig/keymap_data/blender_default.py
+++ b/scripts/presets/keyconfig/keymap_data/blender_default.py
@ -2497,7 +2497,7 @@ def km_nla_generic(_params):
            sidebar_key={"type": 'N', "value": 'PRESS'},
        ),
        ("nla.tweakmode_enter", {"type": 'TAB', "value": 'PRESS'},
-         {"properties": [("use_upper_stack_evaluation", False)]}),
+         {"properties": [("use_upper_stack_evaluation", True)]}),
        ("nla.tweakmode_exit", {"type": 'TAB', "value": 'PRESS'}, None),
        ("nla.tweakmode_enter", {"type": 'TAB', "value": 'PRESS', "shift": True},
         {"properties": [("isolate_action", True)]}),
--- a/scripts/startup/bl_operators/object_quick_effects.py
+++ b/scripts/startup/bl_operators/object_quick_effects.py
@ -9,7 +9,10 @@ from bpy.props import (
    FloatProperty,
    IntProperty,
 )
-from bpy.app.translations import pgettext_tip as tip_
+from bpy.app.translations import (
+    pgettext_tip as tip_,
+    pgettext_data as data_,
+)


 def object_ensure_material(obj, mat_name):
@ -119,7 +122,7 @@ class QuickFur(ObjectModeOperator, Operator):
        noise_group = bpy.data.node_groups["Hair Curves Noise"] if self.use_noise else None
        frizz_group = bpy.data.node_groups["Frizz Hair Curves"] if self.use_frizz else None

-        material = bpy.data.materials.new("Fur Material")
+        material = bpy.data.materials.new(data_("Fur Material"))

        mesh_with_zero_area = False
        mesh_missing_uv_map = False
@ -146,7 +149,7 @@ class QuickFur(ObjectModeOperator, Operator):
            else:
                density = count / area

-            generate_modifier = curves_object.modifiers.new(name="Generate", type='NODES')
+            generate_modifier = curves_object.modifiers.new(name=data_("Generate"), type='NODES')
            generate_modifier.node_group = generate_group
            generate_modifier["Input_2"] = mesh_object
            generate_modifier["Input_18_attribute_name"] = curves.surface_uv_map
@ -155,11 +158,11 @@ class QuickFur(ObjectModeOperator, Operator):
            generate_modifier["Input_22"] = material
            generate_modifier["Input_15"] = density * 0.01

-            radius_modifier = curves_object.modifiers.new(name="Set Hair Curve Profile", type='NODES')
+            radius_modifier = curves_object.modifiers.new(name=data_("Set Hair Curve Profile"), type='NODES')
            radius_modifier.node_group = radius_group
            radius_modifier["Input_3"] = self.radius

-            interpolate_modifier = curves_object.modifiers.new(name="Interpolate Hair Curves", type='NODES')
+            interpolate_modifier = curves_object.modifiers.new(name=data_("Interpolate Hair Curves"), type='NODES')
            interpolate_modifier.node_group = interpolate_group
            interpolate_modifier["Input_2"] = mesh_object
            interpolate_modifier["Input_18_attribute_name"] = curves.surface_uv_map
@ -169,11 +172,11 @@ class QuickFur(ObjectModeOperator, Operator):
            interpolate_modifier["Input_24"] = True

            if noise_group:
-                noise_modifier = curves_object.modifiers.new(name="Hair Curves Noise", type='NODES')
+                noise_modifier = curves_object.modifiers.new(name=data_("Hair Curves Noise"), type='NODES')
                noise_modifier.node_group = noise_group

            if frizz_group:
-                frizz_modifier = curves_object.modifiers.new(name="Frizz Hair Curves", type='NODES')
+                frizz_modifier = curves_object.modifiers.new(name=data_("Frizz Hair Curves"), type='NODES')
                frizz_modifier.node_group = frizz_group

            if self.apply_hair_guides:
--- a/scripts/startup/bl_operators/userpref.py
+++ b/scripts/startup/bl_operators/userpref.py
@ -78,29 +78,30 @@ class PREFERENCES_OT_copy_prev(Operator):

    @classmethod
    def previous_version(cls):
-        # Find config folder from previous version.
        import os
-        version = bpy.app.version
-        version_new = ((version[0] * 100) + version[1])
-        version_old = ((version[0] * 100) + version[1]) - 1
+        # Find config folder from previous version.
+        #
+        # Always allow to load startup data from any release from current major release cycle, and the previous one.

-        # Ensure we only try to copy files from a point release.
-        # The check below ensures the second numbers match.
-        while (version_new % 100) // 10 == (version_old % 100) // 10:
-            version_split = version_old // 100, version_old % 100
-            if os.path.isdir(cls._old_version_path(version_split)):
-                return version_split
-            version_old = version_old - 1
+        # NOTE: This value may need to be updated when the release cycle system is modified.
+        # Here could be `6` in theory (Blender 3.6 LTS), just give it a bit of extra room, such that it does not have to
+        # be updated if there ever exist a 3.7 release e.g.
+        MAX_MINOR_VERSION_FOR_PREVIOUS_MAJOR_LOOKUP = 10

-        # Support loading 2.8x..2.9x startup (any older isn't so useful to load).
-        # NOTE: remove this block for Blender 4.0 and later.
-        if version_old == 299:
-            version_old = 294
-            while version_old >= 280:
-                version_split = version_old // 100, version_old % 100
-                if os.path.isdir(cls._old_version_path(version_split)):
-                    return version_split
-                version_old = version_old - 1
+        version_new = bpy.app.version[:2]
+        version_old = [version_new[0], version_new[1] - 1]
+
+        while True:
+            while version_old[1] >= 0:
+                if os.path.isdir(cls._old_version_path(version_old)):
+                    return tuple(version_old)
+                version_old[1] -= 1
+            if version_new[0] == version_old[0]:
+                # Retry with older major version.
+                version_old[0] -= 1
+                version_old[1] = MAX_MINOR_VERSION_FOR_PREVIOUS_MAJOR_LOOKUP
+            else:
+                break

        return None

--- a/scripts/startup/bl_ui/node_add_menu_geometry.py
+++ b/scripts/startup/bl_ui/node_add_menu_geometry.py
@ -613,6 +613,7 @@ class NODE_MT_category_GEO_VOLUME(Menu):
            layout.separator()
            node_add_menu.add_node_type(layout, "GeometryNodeMeanFilterSDFVolume")
            node_add_menu.add_node_type(layout, "GeometryNodeOffsetSDFVolume")
+            node_add_menu.add_node_type(layout, "GeometryNodeSampleVolume")
            node_add_menu.add_node_type(layout, "GeometryNodeSDFVolumeSphere")
            node_add_menu.add_node_type(layout, "GeometryNodeInputSignedDistance")
        node_add_menu.draw_assets_for_catalog(layout, self.bl_label)
--- a/scripts/startup/bl_ui/properties_data_bone.py
+++ b/scripts/startup/bl_ui/properties_data_bone.py
@ -148,7 +148,7 @@ class BONE_PT_curved(BoneButtonsPanel, Panel):

        col = topcol.column(align=True)
        col.prop(bbone, "bbone_rollin", text="Roll In")
-        col.prop(bbone, "bbone_rollout", text="Out")
+        col.prop(bbone, "bbone_rollout", text="Out", text_ctxt=i18n_contexts.id_armature)
        col.prop(bone, "use_endroll_as_inroll")

        col = topcol.column(align=True)
--- a/scripts/startup/bl_ui/properties_view_layer.py
+++ b/scripts/startup/bl_ui/properties_view_layer.py
@ -147,6 +147,9 @@ class VIEWLAYER_PT_eevee_layer_passes_effects(ViewLayerButtonsPanel, Panel):
        col.prop(view_layer_eevee, "use_pass_bloom", text="Bloom")
        col.active = scene_eevee.use_bloom

+        col = layout.column()
+        col.prop(view_layer_eevee, "use_pass_transparent")
+

 class ViewLayerAOVPanel(ViewLayerButtonsPanel, Panel):
    bl_label = "Shader AOV"
--- a/scripts/startup/bl_ui/space_filebrowser.py
+++ b/scripts/startup/bl_ui/space_filebrowser.py
@ -810,7 +810,7 @@ class ASSETBROWSER_UL_metadata_tags(UIList):
        row = layout.row(align=True)
        # Non-editable entries would show grayed-out, which is bad in this specific case, so switch to mere label.
        if tag.is_property_readonly("name"):
-            row.label(text=tag.name, icon_value=icon)
+            row.label(text=tag.name, icon_value=icon, translate=False)
        else:
            row.prop(tag, "name", text="", emboss=False, icon_value=icon)

--- a/scripts/startup/bl_ui/space_node.py
+++ b/scripts/startup/bl_ui/space_node.py
@ -886,7 +886,8 @@ class NodeTreeInterfacePanel(Panel):
            props = property_row.operator_menu_enum(
                "node.tree_socket_change_type",
                "socket_type",
-                text=active_socket.bl_label if active_socket.bl_label else active_socket.bl_idname,
+                text=(iface_(active_socket.bl_label) if active_socket.bl_label
+                      else iface_(active_socket.bl_idname)),
            )
            props.in_out = in_out

@ -904,10 +905,8 @@ class NodeTreeInterfacePanel(Panel):
                    props = property_row.operator_menu_enum(
                        "node.tree_socket_change_subtype",
                        "socket_subtype",
-                        text=(
-                            active_socket.bl_subtype_label if active_socket.bl_subtype_label else
-                            active_socket.bl_idname
-                        ),
+                        text=(iface_(active_socket.bl_subtype_label) if active_socket.bl_subtype_label
+                              else iface_(active_socket.bl_idname)),
                    )

            layout.use_property_split = True
--- a/scripts/startup/bl_ui/space_sequencer.py
+++ b/scripts/startup/bl_ui/space_sequencer.py
@ -434,6 +434,7 @@ class SEQUENCER_MT_view(Menu):
        if is_sequencer_view:
            layout.operator_context = 'INVOKE_REGION_WIN'
            layout.operator("sequencer.view_all")
+            layout.operator("sequencer.view_frame")
            layout.operator("view2d.zoom_border", text="Zoom")
            layout.prop(st, "use_clamp_view")

--- a/scripts/startup/bl_ui/space_toolsystem_toolbar.py
+++ b/scripts/startup/bl_ui/space_toolsystem_toolbar.py
@ -500,23 +500,21 @@ class _defs_view3d_add:
    # Layout tweaks here would be good to avoid,
    # this shows limits in layout engine, as buttons are using a lot of space.
    @staticmethod
-    def draw_settings_interactive_add(layout, tool, extra):
+    def draw_settings_interactive_add(layout, tool_settings, tool, extra):
        show_extra = False
-        props = tool.operator_properties("view3d.interactive_add")
        if not extra:
            row = layout.row()
            row.label(text="Depth:")
            row = layout.row()
-            row.prop(props, "plane_depth", text="")
+            row.prop(tool_settings, "plane_depth", text="")
            row = layout.row()
            row.label(text="Orientation:")
            row = layout.row()
-            row.prop(props, "plane_orientation", text="")
+            row.prop(tool_settings, "plane_orientation", text="")
            row = layout.row()
-            row.prop(props, "snap_target")
+            row.prop(tool_settings, "snap_elements_tool")

            region_is_header = bpy.context.region.type == 'TOOL_HEADER'
-
            if region_is_header:
                # Don't draw the "extra" popover here as we might have other settings & this should be last.
                show_extra = True
@ -524,9 +522,10 @@ class _defs_view3d_add:
                extra = True

        if extra:
+            props = tool.operator_properties("view3d.interactive_add")
            layout.use_property_split = True
-            layout.row().prop(props, "plane_axis", expand=True)
-            layout.row().prop(props, "plane_axis_auto")
+            layout.row().prop(tool_settings, "plane_axis", expand=True)
+            layout.row().prop(tool_settings, "plane_axis_auto")

            layout.label(text="Base")
            layout.row().prop(props, "plane_origin_base", expand=True)
@ -538,8 +537,8 @@ class _defs_view3d_add:

    @ToolDef.from_fn
    def cube_add():
-        def draw_settings(_context, layout, tool, *, extra=False):
-            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, tool, extra)
+        def draw_settings(context, layout, tool, *, extra=False):
+            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, context.tool_settings, tool, extra)
            if show_extra:
                layout.popover("TOPBAR_PT_tool_settings_extra", text="...")

@ -557,8 +556,8 @@ class _defs_view3d_add:

    @ToolDef.from_fn
    def cone_add():
-        def draw_settings(_context, layout, tool, *, extra=False):
-            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, tool, extra)
+        def draw_settings(context, layout, tool, *, extra=False):
+            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, context.tool_settings, tool, extra)
            if extra:
                return

@ -583,8 +582,8 @@ class _defs_view3d_add:

    @ToolDef.from_fn
    def cylinder_add():
-        def draw_settings(_context, layout, tool, *, extra=False):
-            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, tool, extra)
+        def draw_settings(context, layout, tool, *, extra=False):
+            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, context.tool_settings, tool, extra)
            if extra:
                return

@ -608,8 +607,8 @@ class _defs_view3d_add:

    @ToolDef.from_fn
    def uv_sphere_add():
-        def draw_settings(_context, layout, tool, *, extra=False):
-            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, tool, extra)
+        def draw_settings(context, layout, tool, *, extra=False):
+            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, context.tool_settings, tool, extra)
            if extra:
                return

@ -633,8 +632,8 @@ class _defs_view3d_add:

    @ToolDef.from_fn
    def ico_sphere_add():
-        def draw_settings(_context, layout, tool, *, extra=False):
-            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, tool, extra)
+        def draw_settings(context, layout, tool, *, extra=False):
+            show_extra = _defs_view3d_add.draw_settings_interactive_add(layout, context.tool_settings, tool, extra)
            if extra:
                return

--- a/scripts/startup/nodeitems_builtins.py
+++ b/scripts/startup/nodeitems_builtins.py
@ -191,7 +191,6 @@ shader_node_categories = [
        NodeItem("ShaderNodeBsdfRefraction", poll=object_eevee_cycles_shader_nodes_poll),
        NodeItem("ShaderNodeBsdfGlass", poll=object_eevee_cycles_shader_nodes_poll),
        NodeItem("ShaderNodeBsdfTranslucent", poll=object_eevee_cycles_shader_nodes_poll),
-        NodeItem("ShaderNodeBsdfAnisotropic", poll=object_cycles_shader_nodes_poll),
        NodeItem("ShaderNodeBsdfVelvet", poll=object_cycles_shader_nodes_poll),
        NodeItem("ShaderNodeBsdfToon", poll=object_cycles_shader_nodes_poll),
        NodeItem("ShaderNodeSubsurfaceScattering", poll=object_eevee_cycles_shader_nodes_poll),
--- a/source/blender/CMakeLists.txt
+++ b/source/blender/CMakeLists.txt
@ -61,8 +61,8 @@ set(SRC_DNA_INC
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_pointcache_types.h
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_pointcloud_types.h
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_rigidbody_types.h
-  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_scene_types.h
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_scene_enums.h
+  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_scene_types.h
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_screen_types.h
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_sdna_types.h
  ${CMAKE_CURRENT_SOURCE_DIR}/makesdna/DNA_sequence_types.h
--- a/source/blender/blenkernel/BKE_blender_version.h
+++ b/source/blender/blenkernel/BKE_blender_version.h
@ -17,7 +17,7 @@ extern "C" {
 */

 /* Blender major and minor version. */
-#define BLENDER_VERSION 306
+#define BLENDER_VERSION 400
 /* Blender patch version for bugfix releases. */
 #define BLENDER_VERSION_PATCH 0
 /** Blender release cycle stage: alpha/beta/rc/release. */
@ -25,12 +25,12 @@ extern "C" {

 /* Blender file format version. */
 #define BLENDER_FILE_VERSION BLENDER_VERSION
-#define BLENDER_FILE_SUBVERSION 9
+#define BLENDER_FILE_SUBVERSION 1

 /* Minimum Blender version that supports reading file written with the current
 * version. Older Blender versions will test this and show a warning if the file
 * was written with too new a version. */
-#define BLENDER_FILE_MIN_VERSION 305
+#define BLENDER_FILE_MIN_VERSION 306
 #define BLENDER_FILE_MIN_SUBVERSION 9

 /** User readable version string. */
--- a/source/blender/blenkernel/BKE_collection.h
+++ b/source/blender/blenkernel/BKE_collection.h
@ -322,7 +322,7 @@ void BKE_collection_blend_read_expand(struct BlendExpander *expander,
 void BKE_collection_compat_blend_read_data(struct BlendDataReader *reader,
                                           struct SceneCollection *sc);
 void BKE_collection_compat_blend_read_lib(struct BlendLibReader *reader,
-                                          struct Library *lib,
+                                          struct ID *self_id,
                                          struct SceneCollection *sc);
 void BKE_collection_compat_blend_read_expand(struct BlendExpander *expander,
                                             struct SceneCollection *sc);
--- a/source/blender/blenkernel/BKE_customdata.h
+++ b/source/blender/blenkernel/BKE_customdata.h
@ -314,6 +314,9 @@ void CustomData_free_layers(struct CustomData *data, eCustomDataType type, int t
 * Returns true if a layer with the specified type exists.
 */
 bool CustomData_has_layer(const struct CustomData *data, eCustomDataType type);
+bool CustomData_has_layer_named(const struct CustomData *data,
+                                eCustomDataType type,
+                                const char *name);

 /**
 * Returns the number of layers with this type.
--- a/source/blender/blenkernel/BKE_geometry_set_instances.hh
+++ b/source/blender/blenkernel/BKE_geometry_set_instances.hh
@ -2,8 +2,6 @@

 #pragma once

-#include "BLI_math_matrix_types.hh"
-
 #include "BKE_geometry_set.hh"

 namespace blender::bke {
@ -13,36 +11,4 @@ namespace blender::bke {
 */
 GeometrySet object_get_evaluated_geometry_set(const Object &object);

-/**
- * Used to keep track of a group of instances using the same geometry data.
- */
-struct GeometryInstanceGroup {
-  /**
-   * The geometry set instanced on each of the transforms. The components are not necessarily
-   * owned here. For example, they may be owned by the instanced object. This cannot be a
-   * reference because not all instanced data will necessarily have a #geometry_set_eval.
-   */
-  GeometrySet geometry_set;
-
-  /**
-   * As an optimization to avoid copying, the same geometry set can be associated with multiple
-   * instances. Each instance is stored as a transform matrix here. Again, these must be owned
-   * because they may be transformed from the original data. TODO: Validate that last statement.
-   */
-  Vector<float4x4> transforms;
-};
-
-/**
- * Return flattened vector of the geometry component's recursive instances. I.e. all collection
- * instances and object instances will be expanded into the instances of their geometry components.
- * Even the instances in those geometry components' will be included.
- *
- * \note For convenience (to avoid duplication in the caller), the returned vector also contains
- * the argument geometry set.
- *
- * \note This doesn't extract instances from the "dupli" system for non-geometry-nodes instances.
- */
-void geometry_set_gather_instances(const GeometrySet &geometry_set,
-                                   Vector<GeometryInstanceGroup> &r_instance_groups);
-
 }  // namespace blender::bke
--- a/source/blender/blenkernel/BKE_idprop.h
+++ b/source/blender/blenkernel/BKE_idprop.h
@ -315,7 +315,7 @@ void IDP_BlendReadData_impl(struct BlendDataReader *reader,
                            struct IDProperty **prop,
                            const char *caller_func_id);
 #define IDP_BlendDataRead(reader, prop) IDP_BlendReadData_impl(reader, prop, __func__)
-void IDP_BlendReadLib(struct BlendLibReader *reader, struct Library *lib, struct IDProperty *prop);
+void IDP_BlendReadLib(struct BlendLibReader *reader, struct ID *self_id, struct IDProperty *prop);
 void IDP_BlendReadExpand(struct BlendExpander *expander, struct IDProperty *prop);

 typedef enum eIDPropertyUIDataType {
--- a/source/blender/blenkernel/BKE_image_wrappers.hh
+++ b/source/blender/blenkernel/BKE_image_wrappers.hh
@ -57,13 +57,14 @@ template<typename T, int Channels = 4> struct ImageBufferAccessor {
  {
    if constexpr ((std::is_same_v<T, float4>)) {
      int offset = (coordinate.y * image_buffer.x + coordinate.x) * Channels;
-      return float4(&image_buffer.rect_float[offset]);
+      return float4(&image_buffer.float_buffer.data[offset]);
    }
    if constexpr ((std::is_same_v<T, int>)) {
      int offset = (coordinate.y * image_buffer.x + coordinate.x);
      float4 result;
-      rgba_uchar_to_float(result,
-                          static_cast<uchar *>(static_cast<void *>(&image_buffer.rect[offset])));
+      rgba_uchar_to_float(
+          result,
+          static_cast<uchar *>(static_cast<void *>(&image_buffer.byte_buffer.data[offset])));
      return result;
    }
    return float4();
@ -73,12 +74,13 @@ template<typename T, int Channels = 4> struct ImageBufferAccessor {
  {
    if constexpr ((std::is_same_v<T, float>)) {
      int offset = (coordinate.y * image_buffer.x + coordinate.x) * Channels;
-      copy_v4_v4(&image_buffer.rect_float[offset], new_value);
+      copy_v4_v4(&image_buffer.float_buffer.data[offset], new_value);
    }
    if constexpr ((std::is_same_v<T, int>)) {
      int offset = (coordinate.y * image_buffer.x + coordinate.x);
-      rgba_float_to_uchar(static_cast<uchar *>(static_cast<void *>(&image_buffer.rect[offset])),
-                          new_value);
+      rgba_float_to_uchar(
+          static_cast<uchar *>(static_cast<void *>(&image_buffer.byte_buffer.data[offset])),
+          new_value);
    }
  }
 };
--- a/source/blender/blenkernel/BKE_layer.h
+++ b/source/blender/blenkernel/BKE_layer.h
@ -275,7 +275,7 @@ void BKE_view_layer_blend_write(struct BlendWriter *writer,
                                struct ViewLayer *view_layer);
 void BKE_view_layer_blend_read_data(struct BlendDataReader *reader, struct ViewLayer *view_layer);
 void BKE_view_layer_blend_read_lib(struct BlendLibReader *reader,
-                                   struct Library *lib,
+                                   struct ID *self_id,
                                   struct ViewLayer *view_layer);

 /* iterators */
--- a/source/blender/blenkernel/BKE_lib_query.h
+++ b/source/blender/blenkernel/BKE_lib_query.h
@ -111,12 +111,12 @@ typedef struct LibraryIDLinkCallbackData {
   * 'Real' ID, the one that might be in bmain, only differs from self_id when the later is an
   * embedded one.
   */
-  struct ID *id_owner;
+  struct ID *owner_id;
  /**
   * ID from which the current ID pointer is being processed. It may be an embedded ID like master
   * collection or root node tree.
   */
-  struct ID *id_self;
+  struct ID *self_id;
  struct ID **id_pointer;
  int cb_flag;
 } LibraryIDLinkCallbackData;
@ -259,17 +259,17 @@ void BKE_library_update_ID_link_user(struct ID *id_dst, struct ID *id_src, int c
 int BKE_library_ID_use_ID(struct ID *id_user, struct ID *id_used);

 /**
- * Say whether given \a id_owner may use (in any way) a data-block of \a id_type_used.
+ * Say whether given \a owner_id may use (in any way) a data-block of \a id_type_used.
 *
 * This is a 'simplified' abstract version of #BKE_library_foreach_ID_link() above,
 * quite useful to reduce useless iterations in some cases.
 */
-bool BKE_library_id_can_use_idtype(struct ID *id_owner, short id_type_used);
+bool BKE_library_id_can_use_idtype(struct ID *owner_id, short id_type_used);

 /**
- * Given the id_owner return the type of id_types it can use as a filter_id.
+ * Given the owner_id return the type of id_types it can use as a filter_id.
 */
-uint64_t BKE_library_id_can_use_filter_id(const struct ID *id_owner, const bool include_ui);
+uint64_t BKE_library_id_can_use_filter_id(const struct ID *owner_id, const bool include_ui);

 /**
 * Check whether given ID is used locally (i.e. by another non-linked ID).
--- a/source/blender/blenkernel/BKE_main.h
+++ b/source/blender/blenkernel/BKE_main.h
@ -120,7 +120,18 @@ enum {

 typedef struct Main {
  struct Main *next, *prev;
-  /** The file-path of this blend file, an empty string indicates an unsaved file. */
+  /**
+   * The file-path of this blend file, an empty string indicates an unsaved file.
+   *
+   * \note For the current loaded blend file this path should be absolute & normalized
+   * to prevent redundant leading slashes or current-working-directory relative paths
+   * from causing problems with absolute/relative patch conversion that relies on this being
+   * an absolute path. See #BLI_path_canonicalize_native.
+   *
+   * This rule is not strictly enforced as in some cases loading a #Main is performed
+   * to read data temporarily (preferences & startup) for e.g.
+   * where the `filepath` is not persistent or used as a basis for other paths.
+   */
  char filepath[1024];               /* 1024 = FILE_MAX */
  short versionfile, subversionfile; /* see BLENDER_FILE_VERSION, BLENDER_FILE_SUBVERSION */
  short minversionfile, minsubversionfile;
--- a/source/blender/blenkernel/BKE_mesh_legacy_convert.h
+++ b/source/blender/blenkernel/BKE_mesh_legacy_convert.h
@ -30,107 +30,51 @@ void BKE_mesh_legacy_convert_uvs_to_struct(
    blender::Vector<CustomDataLayer, 16> &loop_layers_to_write);
 void BKE_mesh_legacy_convert_uvs_to_generic(Mesh *mesh);

-/**
- * Move face sets to the legacy type from a generic type.
- */
-void BKE_mesh_legacy_face_set_from_generic(
-    blender::MutableSpan<CustomDataLayer> poly_layers_to_write);
 /**
 * Copy face sets to the generic data type from the legacy type.
 */
 void BKE_mesh_legacy_face_set_to_generic(struct Mesh *mesh);

-/**
- * Copy edge creases from a separate layer into edges.
- */
-void BKE_mesh_legacy_edge_crease_from_layers(struct Mesh *mesh);
 /**
 * Copy edge creases from edges to a separate layer.
 */
 void BKE_mesh_legacy_edge_crease_to_layers(struct Mesh *mesh);

-/**
- * Copy bevel weights from separate layers into vertices and edges.
- */
-void BKE_mesh_legacy_bevel_weight_from_layers(struct Mesh *mesh);
 /**
 * Copy bevel weights from vertices and edges to separate layers.
 */
 void BKE_mesh_legacy_bevel_weight_to_layers(struct Mesh *mesh);

-/**
- * Convert the hidden element attributes to the old flag format for writing.
- */
-void BKE_mesh_legacy_convert_hide_layers_to_flags(struct Mesh *mesh,
-                                                  blender::MutableSpan<MPoly> legacy_polys);
 /**
 * Convert the old hide flags (#ME_HIDE) to the hidden element attribute for reading.
 * Only add the attributes when there are any elements in each domain hidden.
 */
 void BKE_mesh_legacy_convert_flags_to_hide_layers(struct Mesh *mesh);

-/**
- * Convert the selected element attributes to the old flag format for writing.
- */
-void BKE_mesh_legacy_convert_selection_layers_to_flags(struct Mesh *mesh,
-                                                       blender::MutableSpan<MPoly> legacy_polys);
 /**
 * Convert the old selection flags (#SELECT/#ME_FACE_SEL) to the selected element attribute for
 * reading. Only add the attributes when there are any elements in each domain selected.
 */
 void BKE_mesh_legacy_convert_flags_to_selection_layers(struct Mesh *mesh);

-/**
- * Move material indices from a generic attribute to #MPoly.
- */
-void BKE_mesh_legacy_convert_material_indices_to_mpoly(struct Mesh *mesh,
-                                                       blender::MutableSpan<MPoly> legacy_polys);
 /**
 * Move material indices from the #MPoly struct to a generic attributes.
 * Only add the attribute when the indices are not all zero.
 */
 void BKE_mesh_legacy_convert_mpoly_to_material_indices(struct Mesh *mesh);

-/** Convert from runtime loose edge cache to legacy edge flag. */
-void BKE_mesh_legacy_convert_loose_edges_to_flag(struct Mesh *mesh);
-
 void BKE_mesh_legacy_attribute_flags_to_strings(struct Mesh *mesh);
-void BKE_mesh_legacy_attribute_strings_to_flags(struct Mesh *mesh);

-void BKE_mesh_legacy_sharp_faces_to_flags(struct Mesh *mesh,
-                                          blender::MutableSpan<MPoly> legacy_polys);
 void BKE_mesh_legacy_sharp_faces_from_flags(struct Mesh *mesh);

-void BKE_mesh_legacy_sharp_edges_to_flags(struct Mesh *mesh);
 void BKE_mesh_legacy_sharp_edges_from_flags(struct Mesh *mesh);

-void BKE_mesh_legacy_uv_seam_to_flags(struct Mesh *mesh);
 void BKE_mesh_legacy_uv_seam_from_flags(struct Mesh *mesh);

-struct MVert *BKE_mesh_legacy_convert_positions_to_verts(
-    Mesh *mesh,
-    blender::ResourceScope &temp_arrays_for_convert,
-    blender::Vector<CustomDataLayer, 16> &vert_layers_to_write);
-
 void BKE_mesh_legacy_convert_verts_to_positions(Mesh *mesh);

-MEdge *BKE_mesh_legacy_convert_edges_to_medge(
-    Mesh *mesh,
-    blender::ResourceScope &temp_arrays_for_convert,
-    blender::Vector<CustomDataLayer, 16> &edge_layers_to_write);
-
 void BKE_mesh_legacy_convert_edges_to_generic(Mesh *mesh);

-struct MLoop *BKE_mesh_legacy_convert_corners_to_loops(
-    Mesh *mesh,
-    blender::ResourceScope &temp_arrays_for_convert,
-    blender::Vector<CustomDataLayer, 16> &loop_layers_to_write);
-
-blender::MutableSpan<MPoly> BKE_mesh_legacy_convert_offsets_to_polys(
-    const Mesh *mesh,
-    blender::ResourceScope &temp_arrays_for_convert,
-    blender::Vector<CustomDataLayer, 16> &poly_layers_to_write);
-
 void BKE_mesh_legacy_convert_polys_to_offsets(Mesh *mesh);

 void BKE_mesh_legacy_convert_loops_to_corners(struct Mesh *mesh);
--- a/source/blender/blenkernel/BKE_mesh_sample.hh
+++ b/source/blender/blenkernel/BKE_mesh_sample.hh
@ -35,6 +35,14 @@ void sample_point_attribute(Span<int> corner_verts,
                            IndexMask mask,
                            GMutableSpan dst);

+void sample_point_normals(Span<int> corner_verts,
+                          Span<MLoopTri> looptris,
+                          Span<int> looptri_indices,
+                          Span<float3> bary_coords,
+                          Span<float3> src,
+                          IndexMask mask,
+                          MutableSpan<float3> dst);
+
 void sample_corner_attribute(Span<MLoopTri> looptris,
                             Span<int> looptri_indices,
                             Span<float3> bary_coords,
--- a/source/blender/blenkernel/BKE_node.h
+++ b/source/blender/blenkernel/BKE_node.h
@ -554,8 +554,10 @@ struct bNodeSocket *ntreeAddSocketInterface(struct bNodeTree *ntree,
 * \{ */

 struct bNodeType *nodeTypeFind(const char *idname);
+const char *nodeTypeFindAlias(const char *idname);
 void nodeRegisterType(struct bNodeType *ntype);
 void nodeUnregisterType(struct bNodeType *ntype);
+void nodeRegisterAlias(struct bNodeType *nt, const char *alias);
 struct GHashIterator *nodeTypeGetIterator(void);

 /* Helper macros for iterating over node types. */
@ -893,9 +895,9 @@ void BKE_nodetree_remove_layer_n(struct bNodeTree *ntree, struct Scene *scene, i
 #define SH_NODE_MIX_SHADER 128
 #define SH_NODE_ATTRIBUTE 129
 #define SH_NODE_BACKGROUND 130
-#define SH_NODE_BSDF_ANISOTROPIC 131
+#define SH_NODE_BSDF_GLOSSY 131
 #define SH_NODE_BSDF_DIFFUSE 132
-#define SH_NODE_BSDF_GLOSSY 133
+#define SH_NODE_BSDF_GLOSSY_LEGACY 133
 #define SH_NODE_BSDF_GLASS 134
 #define SH_NODE_BSDF_TRANSLUCENT 137
 #define SH_NODE_BSDF_TRANSPARENT 138
@ -1304,6 +1306,7 @@ void BKE_nodetree_remove_layer_n(struct bNodeTree *ntree, struct Scene *scene, i
 #define GEO_NODE_SIMULATION_INPUT 2100
 #define GEO_NODE_SIMULATION_OUTPUT 2101
 #define GEO_NODE_INPUT_SIGNED_DISTANCE 2102
+#define GEO_NODE_SAMPLE_VOLUME 2103

 /** \} */

--- a/source/blender/blenkernel/BKE_node.hh
+++ b/source/blender/blenkernel/BKE_node.hh
@ -39,8 +39,6 @@ bNodeTree *ntreeCopyTree(Main *bmain, const bNodeTree *ntree);

 void ntreeFreeLocalNode(bNodeTree *ntree, bNode *node);

-bNode *ntreeFindType(bNodeTree *ntree, int type);
-
 void ntreeUpdateAllNew(Main *main);

 void ntreeNodeFlagSet(const bNodeTree *ntree, int flag, bool enable);
@ -215,44 +213,38 @@ void nodeParentsIter(bNode *node, bool (*callback)(bNode *, void *), void *userd
 * A dangling reroute node is a reroute node that does *not* have a "data source", i.e. no
 * non-reroute node is connected to its input.
 */
-bool nodeIsDanglingReroute(const struct bNodeTree *ntree, const struct bNode *node);
+bool nodeIsDanglingReroute(const bNodeTree *ntree, const bNode *node);

-struct bNodeLink *nodeFindLink(struct bNodeTree *ntree,
-                               const struct bNodeSocket *from,
-                               const struct bNodeSocket *to);
-
-struct bNode *nodeGetActivePaintCanvas(struct bNodeTree *ntree);
+bNode *nodeGetActivePaintCanvas(bNodeTree *ntree);

 /**
 * \brief Does the given node supports the sub active flag.
 *
 * \param sub_active: The active flag to check. #NODE_ACTIVE_TEXTURE / #NODE_ACTIVE_PAINT_CANVAS.
 */
-bool nodeSupportsActiveFlag(const struct bNode *node, int sub_active);
+bool nodeSupportsActiveFlag(const bNode *node, int sub_active);

-void nodeSetSocketAvailability(struct bNodeTree *ntree,
-                               struct bNodeSocket *sock,
-                               bool is_available);
+void nodeSetSocketAvailability(bNodeTree *ntree, bNodeSocket *sock, bool is_available);

 /**
 * If the node implements a `declare` function, this function makes sure that `node->declaration`
 * is up to date. It is expected that the sockets of the node are up to date already.
 */
-bool nodeDeclarationEnsure(struct bNodeTree *ntree, struct bNode *node);
+bool nodeDeclarationEnsure(bNodeTree *ntree, bNode *node);

 /**
 * Just update `node->declaration` if necessary. This can also be called on nodes that may not be
 * up to date (e.g. because the need versioning or are dynamic).
 */
-bool nodeDeclarationEnsureOnOutdatedNode(struct bNodeTree *ntree, struct bNode *node);
+bool nodeDeclarationEnsureOnOutdatedNode(bNodeTree *ntree, bNode *node);

 /**
 * Update `socket->declaration` for all sockets in the node. This assumes that the node declaration
 * and sockets are up to date already.
 */
-void nodeSocketDeclarationsUpdate(struct bNode *node);
+void nodeSocketDeclarationsUpdate(bNode *node);

-typedef GHashIterator bNodeInstanceHashIterator;
+using bNodeInstanceHashIterator = GHashIterator;

 BLI_INLINE bNodeInstanceHashIterator *node_instance_hash_iterator_new(bNodeInstanceHash *hash)
 {
@ -316,6 +308,7 @@ void node_preview_merge_tree(bNodeTree *to_ntree, bNodeTree *from_ntree, bool re
 /* -------------------------------------------------------------------- */
 /** \name Node Type Access
 * \{ */
+
 void nodeLabel(const bNodeTree *ntree, const bNode *node, char *label, int maxlen);

 /**
@ -347,7 +340,7 @@ void node_type_size_preset(bNodeType *ntype, eNodeSizePreset size);
 /** \name Node Generic Functions
 * \{ */

-bool node_is_connected_to_output(const struct bNodeTree *ntree, const struct bNode *node);
+bool node_is_connected_to_output(const bNodeTree *ntree, const bNode *node);

 bNodeSocket *node_find_enabled_socket(bNode &node, eNodeSocketInOut in_out, StringRef name);

--- a/source/blender/blenkernel/BKE_pbvh_pixels.hh
+++ b/source/blender/blenkernel/BKE_pbvh_pixels.hh
@ -346,7 +346,7 @@ struct CopyPixelTile {

  void copy_pixels(ImBuf &tile_buffer, IndexRange group_range) const
  {
-    if (tile_buffer.rect_float) {
+    if (tile_buffer.float_buffer.data) {
      image::ImageBufferAccessor<float4> accessor(tile_buffer);
      copy_pixels<float4>(accessor, group_range);
    }
--- a/source/blender/blenkernel/BKE_sound.h
+++ b/source/blender/blenkernel/BKE_sound.h
@ -154,6 +154,8 @@ void BKE_sound_set_scene_sound_volume(void *handle, float volume, char animated)

 void BKE_sound_set_scene_sound_pitch(void *handle, float pitch, char animated);

+void BKE_sound_set_scene_sound_pitch_at_frame(void *handle, int frame, float pitch, char animated);
+
 void BKE_sound_set_scene_sound_pitch_constant_range(void *handle,
                                                    int frame_start,
                                                    int frame_end,
--- a/source/blender/blenkernel/BKE_text.h
+++ b/source/blender/blenkernel/BKE_text.h
@ -94,6 +94,9 @@ void txt_sel_clear(struct Text *text);
 void txt_sel_line(struct Text *text);
 void txt_sel_set(struct Text *text, int startl, int startc, int endl, int endc);
 char *txt_sel_to_buf(const struct Text *text, size_t *r_buf_strlen);
+/**
+ * \param in_buffer: UTF8 encoded text, invalid UTF8 byte-sequences are handled gracefully.
+ */
 void txt_insert_buf(struct Text *text, const char *in_buffer, int in_buffer_len)
    ATTR_NONNULL(1, 2);
 void txt_split_curline(struct Text *text);
--- a/source/blender/blenkernel/BKE_viewer_path.h
+++ b/source/blender/blenkernel/BKE_viewer_path.h
@ -40,7 +40,7 @@ bool BKE_viewer_path_equal(const ViewerPath *a, const ViewerPath *b);
 void BKE_viewer_path_blend_write(struct BlendWriter *writer, const ViewerPath *viewer_path);
 void BKE_viewer_path_blend_read_data(struct BlendDataReader *reader, ViewerPath *viewer_path);
 void BKE_viewer_path_blend_read_lib(struct BlendLibReader *reader,
-                                    struct Library *lib,
+                                    struct ID *self_id,
                                    ViewerPath *viewer_path);
 void BKE_viewer_path_foreach_id(struct LibraryForeachIDData *data, ViewerPath *viewer_path);
 void BKE_viewer_path_id_remap(ViewerPath *viewer_path, const struct IDRemapper *mappings);
--- a/source/blender/blenkernel/CMakeLists.txt
+++ b/source/blender/blenkernel/CMakeLists.txt
@ -443,6 +443,7 @@ set(SRC
  BKE_multires.h
  BKE_nla.h
  BKE_node.h
+  BKE_node.hh
  BKE_node_runtime.hh
  BKE_node_tree_update.h
  BKE_node_tree_zones.hh
--- a/source/blender/blenkernel/intern/action.c
+++ b/source/blender/blenkernel/intern/action.c
@ -222,7 +222,7 @@ static void action_blend_read_data(BlendDataReader *reader, ID *id)
 static void blend_read_lib_constraint_channels(BlendLibReader *reader, ID *id, ListBase *chanbase)
 {
  LISTBASE_FOREACH (bConstraintChannel *, chan, chanbase) {
-    BLO_read_id_address(reader, id->lib, &chan->ipo);
+    BLO_read_id_address(reader, id, &chan->ipo);
  }
 }

@ -232,16 +232,16 @@ static void action_blend_read_lib(BlendLibReader *reader, ID *id)

  /* XXX deprecated - old animation system <<< */
  LISTBASE_FOREACH (bActionChannel *, chan, &act->chanbase) {
-    BLO_read_id_address(reader, act->id.lib, &chan->ipo);
+    BLO_read_id_address(reader, id, &chan->ipo);
    blend_read_lib_constraint_channels(reader, &act->id, &chan->constraintChannels);
  }
  /* >>> XXX deprecated - old animation system */

-  BKE_fcurve_blend_read_lib(reader, &act->id, &act->curves);
+  BKE_fcurve_blend_read_lib(reader, id, &act->curves);

  LISTBASE_FOREACH (TimeMarker *, marker, &act->markers) {
    if (marker->camera) {
-      BLO_read_id_address(reader, act->id.lib, &marker->camera);
+      BLO_read_id_address(reader, id, &marker->camera);
    }
  }
 }
@ -1952,9 +1952,9 @@ void BKE_pose_blend_read_lib(BlendLibReader *reader, Object *ob, bPose *pose)

    pchan->bone = BKE_armature_find_bone_name(arm, pchan->name);

-    IDP_BlendReadLib(reader, ob->id.lib, pchan->prop);
+    IDP_BlendReadLib(reader, &ob->id, pchan->prop);

-    BLO_read_id_address(reader, ob->id.lib, &pchan->custom);
+    BLO_read_id_address(reader, &ob->id, &pchan->custom);
    if (UNLIKELY(pchan->bone == NULL)) {
      rebuild = true;
    }
--- a/source/blender/blenkernel/intern/anim_data.c
+++ b/source/blender/blenkernel/intern/anim_data.c
@ -1479,8 +1479,8 @@ void BKE_animdata_blend_read_lib(BlendLibReader *reader, ID *id, AnimData *adt)
  }

  /* link action data */
-  BLO_read_id_address(reader, id->lib, &adt->action);
-  BLO_read_id_address(reader, id->lib, &adt->tmpact);
+  BLO_read_id_address(reader, id, &adt->action);
+  BLO_read_id_address(reader, id, &adt->tmpact);

  /* link drivers */
  BKE_fcurve_blend_read_lib(reader, id, &adt->drivers);
--- a/source/blender/blenkernel/intern/anim_sys.c
+++ b/source/blender/blenkernel/intern/anim_sys.c
@ -329,7 +329,7 @@ void BKE_keyingsets_blend_read_lib(BlendLibReader *reader, ID *id, ListBase *lis
 {
  LISTBASE_FOREACH (KeyingSet *, ks, list) {
    LISTBASE_FOREACH (KS_Path *, ksp, &ks->paths) {
-      BLO_read_id_address(reader, id->lib, &ksp->id);
+      BLO_read_id_address(reader, id, &ksp->id);
    }
  }
 }
--- a/source/blender/blenkernel/intern/appdir.c
+++ b/source/blender/blenkernel/intern/appdir.c
@ -897,8 +897,7 @@ void BKE_appdir_program_path_init(const char *argv0)
   * Otherwise other methods of detecting the binary that override this argument
   * which must point to the Python module for data-files to be detected. */
  STRNCPY(g_app.program_filepath, argv0);
-  BLI_path_abs_from_cwd(g_app.program_filepath, sizeof(g_app.program_filepath));
-  BLI_path_normalize_native(g_app.program_filepath);
+  BLI_path_canonicalize_native(g_app.program_filepath, sizeof(g_app.program_filepath));

  if (g_app.program_dirname[0] == '\0') {
    /* First time initializing, the file binary path isn't valid from a Python module.
--- a/source/blender/blenkernel/intern/armature.c
+++ b/source/blender/blenkernel/intern/armature.c
@ -261,12 +261,12 @@ static void armature_blend_read_data(BlendDataReader *reader, ID *id)
  BKE_armature_bone_hash_make(arm);
 }

-static void lib_link_bones(BlendLibReader *reader, Library *lib, Bone *bone)
+static void lib_link_bones(BlendLibReader *reader, ID *self_id, Bone *bone)
 {
-  IDP_BlendReadLib(reader, lib, bone->prop);
+  IDP_BlendReadLib(reader, self_id, bone->prop);

  LISTBASE_FOREACH (Bone *, curbone, &bone->childbase) {
-    lib_link_bones(reader, lib, curbone);
+    lib_link_bones(reader, self_id, curbone);
  }
 }

@ -274,7 +274,7 @@ static void armature_blend_read_lib(BlendLibReader *reader, ID *id)
 {
  bArmature *arm = (bArmature *)id;
  LISTBASE_FOREACH (Bone *, curbone, &arm->bonebase) {
-    lib_link_bones(reader, id->lib, curbone);
+    lib_link_bones(reader, id, curbone);
  }
 }

--- a/source/blender/blenkernel/intern/attribute_access.cc
+++ b/source/blender/blenkernel/intern/attribute_access.cc
@ -446,7 +446,7 @@ bool BuiltinCustomDataLayerProvider::try_create(void *owner,

  const int element_num = custom_data_access_.get_element_num(owner);
  if (stored_as_named_attribute_) {
-    if (CustomData_get_layer_named(custom_data, data_type_, name_.c_str())) {
+    if (CustomData_has_layer_named(custom_data, data_type_, name_.c_str())) {
      /* Exists already. */
      return false;
    }
@ -469,7 +469,7 @@ bool BuiltinCustomDataLayerProvider::exists(const void *owner) const
    return false;
  }
  if (stored_as_named_attribute_) {
-    return CustomData_get_layer_named(custom_data, stored_type_, name_.c_str()) != nullptr;
+    return CustomData_has_layer_named(custom_data, stored_type_, name_.c_str());
  }
  return CustomData_get_layer(custom_data, stored_type_) != nullptr;
 }
--- a/source/blender/blenkernel/intern/blendfile_link_append.c
+++ b/source/blender/blenkernel/intern/blendfile_link_append.c
@ -952,7 +952,7 @@ static int foreach_libblock_link_append_callback(LibraryIDLinkCallbackData *cb_d
     * meshes for shape keys e.g.), or this is an unsupported case (two shape-keys depending on
     * each-other need to be also 'linked' in by their respective meshes, independent shape-keys
     * are not allowed). ref #96048. */
-    if (id != cb_data->id_self && BKE_idtype_idcode_is_linkable(GS(cb_data->id_self->name))) {
+    if (id != cb_data->self_id && BKE_idtype_idcode_is_linkable(GS(cb_data->self_id->name))) {
      BKE_library_foreach_ID_link(
          cb_data->bmain, id, foreach_libblock_link_append_callback, data, IDWALK_NOP);
    }
@ -972,7 +972,7 @@ static int foreach_libblock_link_append_callback(LibraryIDLinkCallbackData *cb_d
  const bool do_recursive = (data->lapp_context->params->flag & BLO_LIBLINK_APPEND_RECURSIVE) !=
                                0 ||
                            do_link;
-  if (!do_recursive && cb_data->id_owner->lib != id->lib) {
+  if (!do_recursive && cb_data->owner_id->lib != id->lib) {
    return IDWALK_RET_NOP;
  }

--- a/source/blender/blenkernel/intern/bpath.c
+++ b/source/blender/blenkernel/intern/bpath.c
@ -513,7 +513,14 @@ static bool relative_convert_foreach_path_cb(BPathForeachPathData *bpath_data,
    data->count_changed++;
  }
  else {
-    BKE_reportf(data->reports, RPT_WARNING, "Path '%s' cannot be made relative", path_src);
+    const char *type_name = BKE_idtype_get_info_from_id(bpath_data->owner_id)->name;
+    const char *id_name = bpath_data->owner_id->name + 2;
+    BKE_reportf(data->reports,
+                RPT_WARNING,
+                "Path '%s' cannot be made relative for %s '%s'",
+                path_src,
+                type_name,
+                id_name);
    data->count_failed++;
  }
  return true;
@ -537,7 +544,14 @@ static bool absolute_convert_foreach_path_cb(BPathForeachPathData *bpath_data,
    data->count_changed++;
  }
  else {
-    BKE_reportf(data->reports, RPT_WARNING, "Path '%s' cannot be made absolute", path_src);
+    const char *type_name = BKE_idtype_get_info_from_id(bpath_data->owner_id)->name;
+    const char *id_name = bpath_data->owner_id->name + 2;
+    BKE_reportf(data->reports,
+                RPT_WARNING,
+                "Path '%s' cannot be made absolute for %s '%s'",
+                path_src,
+                type_name,
+                id_name);
    data->count_failed++;
  }
  return true;
--- a/source/blender/blenkernel/intern/brush.cc
+++ b/source/blender/blenkernel/intern/brush.cc
@ -361,16 +361,16 @@ static void brush_blend_read_lib(BlendLibReader *reader, ID *id)
  Brush *brush = (Brush *)id;

  /* brush->(mask_)mtex.obj is ignored on purpose? */
-  BLO_read_id_address(reader, brush->id.lib, &brush->mtex.tex);
-  BLO_read_id_address(reader, brush->id.lib, &brush->mask_mtex.tex);
-  BLO_read_id_address(reader, brush->id.lib, &brush->clone.image);
-  BLO_read_id_address(reader, brush->id.lib, &brush->toggle_brush);
-  BLO_read_id_address(reader, brush->id.lib, &brush->paint_curve);
+  BLO_read_id_address(reader, id, &brush->mtex.tex);
+  BLO_read_id_address(reader, id, &brush->mask_mtex.tex);
+  BLO_read_id_address(reader, id, &brush->clone.image);
+  BLO_read_id_address(reader, id, &brush->toggle_brush);
+  BLO_read_id_address(reader, id, &brush->paint_curve);

  /* link default grease pencil palette */
  if (brush->gpencil_settings != nullptr) {
    if (brush->gpencil_settings->flag & GP_BRUSH_MATERIAL_PINNED) {
-      BLO_read_id_address(reader, brush->id.lib, &brush->gpencil_settings->material);
+      BLO_read_id_address(reader, id, &brush->gpencil_settings->material);

      if (!brush->gpencil_settings->material) {
        brush->gpencil_settings->flag &= ~GP_BRUSH_MATERIAL_PINNED;
@ -379,7 +379,7 @@ static void brush_blend_read_lib(BlendLibReader *reader, ID *id)
    else {
      brush->gpencil_settings->material = nullptr;
    }
-    BLO_read_id_address(reader, brush->id.lib, &brush->gpencil_settings->material_alt);
+    BLO_read_id_address(reader, id, &brush->gpencil_settings->material_alt);
  }
 }

@ -399,10 +399,12 @@ static void brush_blend_read_expand(BlendExpander *expander, ID *id)
 static int brush_undo_preserve_cb(LibraryIDLinkCallbackData *cb_data)
 {
  BlendLibReader *reader = (BlendLibReader *)cb_data->user_data;
+  ID *self_id = cb_data->self_id;
  ID *id_old = *cb_data->id_pointer;
  /* Old data has not been remapped to new values of the pointers, if we want to keep the old
   * pointer here we need its new address. */
-  ID *id_old_new = id_old != nullptr ? BLO_read_get_new_id_address(reader, id_old->lib, id_old) :
+  ID *id_old_new = id_old != nullptr ? BLO_read_get_new_id_address(
+                                           reader, self_id, ID_IS_LINKED(self_id), id_old) :
                                       nullptr;
  BLI_assert(id_old_new == nullptr || ELEM(id_old, id_old_new, id_old_new->orig_id));
  if (cb_data->cb_flag & IDWALK_CB_USER) {
@ -2622,18 +2624,22 @@ struct ImBuf *BKE_brush_gen_radial_control_imbuf(Brush *br, bool secondary, bool
  int half = side / 2;

  BKE_curvemapping_init(br->curve);
-  im->rect_float = (float *)MEM_callocN(sizeof(float) * side * side, "radial control rect");
+
+  float *rect_float = (float *)MEM_callocN(sizeof(float) * side * side, "radial control rect");
+  IMB_assign_float_buffer(im, rect_float, IB_DO_NOT_TAKE_OWNERSHIP);
+
  im->x = im->y = side;

-  const bool have_texture = brush_gen_texture(br, side, secondary, im->rect_float);
+  const bool have_texture = brush_gen_texture(br, side, secondary, im->float_buffer.data);

  if (display_gradient || have_texture) {
    for (int i = 0; i < side; i++) {
      for (int j = 0; j < side; j++) {
        const float magn = sqrtf(pow2f(i - half) + pow2f(j - half));
        const float strength = BKE_brush_curve_strength_clamped(br, magn, half);
-        im->rect_float[i * side + j] = (have_texture) ? im->rect_float[i * side + j] * strength :
-                                                        strength;
+        im->float_buffer.data[i * side + j] = (have_texture) ?
+                                                  im->float_buffer.data[i * side + j] * strength :
+                                                  strength;
      }
    }
  }
--- a/source/blender/blenkernel/intern/camera.c
+++ b/source/blender/blenkernel/intern/camera.c
@ -142,14 +142,14 @@ static void camera_blend_read_data(BlendDataReader *reader, ID *id)
 static void camera_blend_read_lib(BlendLibReader *reader, ID *id)
 {
  Camera *ca = (Camera *)id;
-  BLO_read_id_address(reader, ca->id.lib, &ca->ipo); /* deprecated, for versioning */
+  BLO_read_id_address(reader, id, &ca->ipo); /* deprecated, for versioning */

-  BLO_read_id_address(reader, ca->id.lib, &ca->dof_ob); /* deprecated, for versioning */
-  BLO_read_id_address(reader, ca->id.lib, &ca->dof.focus_object);
+  BLO_read_id_address(reader, id, &ca->dof_ob); /* deprecated, for versioning */
+  BLO_read_id_address(reader, id, &ca->dof.focus_object);

  LISTBASE_FOREACH (CameraBGImage *, bgpic, &ca->bg_images) {
-    BLO_read_id_address(reader, ca->id.lib, &bgpic->ima);
-    BLO_read_id_address(reader, ca->id.lib, &bgpic->clip);
+    BLO_read_id_address(reader, id, &bgpic->ima);
+    BLO_read_id_address(reader, id, &bgpic->clip);
  }
 }

--- a/source/blender/blenkernel/intern/collection.c
+++ b/source/blender/blenkernel/intern/collection.c
@ -325,11 +325,11 @@ static void collection_blend_read_data(BlendDataReader *reader, ID *id)
  BKE_collection_blend_read_data(reader, collection, NULL);
 }

-static void lib_link_collection_data(BlendLibReader *reader, Library *lib, Collection *collection)
+static void lib_link_collection_data(BlendLibReader *reader, ID *self_id, Collection *collection)
 {
  BLI_assert(collection->runtime.gobject_hash == NULL);
  LISTBASE_FOREACH_MUTABLE (CollectionObject *, cob, &collection->gobject) {
-    BLO_read_id_address(reader, lib, &cob->ob);
+    BLO_read_id_address(reader, self_id, &cob->ob);

    if (cob->ob == NULL) {
      BLI_freelinkN(&collection->gobject, cob);
@ -337,22 +337,20 @@ static void lib_link_collection_data(BlendLibReader *reader, Library *lib, Colle
  }

  LISTBASE_FOREACH (CollectionChild *, child, &collection->children) {
-    BLO_read_id_address(reader, lib, &child->collection);
+    BLO_read_id_address(reader, self_id, &child->collection);
  }
 }

 #ifdef USE_COLLECTION_COMPAT_28
-void BKE_collection_compat_blend_read_lib(BlendLibReader *reader,
-                                          Library *lib,
-                                          SceneCollection *sc)
+void BKE_collection_compat_blend_read_lib(BlendLibReader *reader, ID *self_id, SceneCollection *sc)
 {
  LISTBASE_FOREACH (LinkData *, link, &sc->objects) {
-    BLO_read_id_address(reader, lib, &link->data);
+    BLO_read_id_address(reader, self_id, &link->data);
    BLI_assert(link->data);
  }

  LISTBASE_FOREACH (SceneCollection *, nsc, &sc->scene_collections) {
-    BKE_collection_compat_blend_read_lib(reader, lib, nsc);
+    BKE_collection_compat_blend_read_lib(reader, self_id, nsc);
  }
 }
 #endif
@ -361,15 +359,15 @@ void BKE_collection_blend_read_lib(BlendLibReader *reader, Collection *collectio
 {
 #ifdef USE_COLLECTION_COMPAT_28
  if (collection->collection) {
-    BKE_collection_compat_blend_read_lib(reader, collection->id.lib, collection->collection);
+    BKE_collection_compat_blend_read_lib(reader, &collection->id, collection->collection);
  }

  if (collection->view_layer) {
-    BKE_view_layer_blend_read_lib(reader, collection->id.lib, collection->view_layer);
+    BKE_view_layer_blend_read_lib(reader, &collection->id, collection->view_layer);
  }
 #endif

-  lib_link_collection_data(reader, collection->id.lib, collection);
+  lib_link_collection_data(reader, &collection->id, collection);
 }

 static void collection_blend_read_lib(BlendLibReader *reader, ID *id)
--- a/source/blender/blenkernel/intern/colortools.c
+++ b/source/blender/blenkernel/intern/colortools.c
@ -1382,11 +1382,11 @@ void BKE_histogram_update_sample_line(Histogram *hist,
  hist->xmax = 1.0f;
  /* hist->ymax = 1.0f; */ /* now do this on the operator _only_ */

-  if (ibuf->rect == NULL && ibuf->rect_float == NULL) {
+  if (ibuf->byte_buffer.data == NULL && ibuf->float_buffer.data == NULL) {
    return;
  }

-  if (ibuf->rect_float) {
+  if (ibuf->float_buffer.data) {
    cm_processor = IMB_colormanagement_display_processor_new(view_settings, display_settings);
  }

@ -1399,9 +1399,9 @@ void BKE_histogram_update_sample_line(Histogram *hist,
          0.0f;
    }
    else {
-      if (ibuf->rect_float) {
+      if (ibuf->float_buffer.data) {
        float rgba[4];
-        fp = (ibuf->rect_float + (ibuf->channels) * (y * ibuf->x + x));
+        fp = (ibuf->float_buffer.data + (ibuf->channels) * (y * ibuf->x + x));

        switch (ibuf->channels) {
          case 4:
@ -1431,8 +1431,8 @@ void BKE_histogram_update_sample_line(Histogram *hist,
        hist->data_b[i] = rgba[2];
        hist->data_a[i] = rgba[3];
      }
-      else if (ibuf->rect) {
-        cp = (uchar *)(ibuf->rect + y * ibuf->x + x);
+      else if (ibuf->byte_buffer.data) {
+        cp = ibuf->byte_buffer.data + 4 * (y * ibuf->x + x);
        hist->data_luma[i] = (float)IMB_colormanagement_get_luminance_byte(cp) / 255.0f;
        hist->data_r[i] = (float)cp[0] / 255.0f;
        hist->data_g[i] = (float)cp[1] / 255.0f;
@ -1492,10 +1492,10 @@ static void scopes_update_cb(void *__restrict userdata,
  const int savedlines = y / rows_per_sample_line;
  const bool do_sample_line = (savedlines < scopes->sample_lines) &&
                              (y % rows_per_sample_line) == 0;
-  const bool is_float = (ibuf->rect_float != NULL);
+  const bool is_float = (ibuf->float_buffer.data != NULL);

  if (is_float) {
-    rf = ibuf->rect_float + ((size_t)y) * ibuf->x * ibuf->channels;
+    rf = ibuf->float_buffer.data + ((size_t)y) * ibuf->x * ibuf->channels;
  }
  else {
    rc = display_buffer + ((size_t)y) * ibuf->x * ibuf->channels;
@ -1616,7 +1616,7 @@ void BKE_scopes_update(Scopes *scopes,
  void *cache_handle = NULL;
  struct ColormanageProcessor *cm_processor = NULL;

-  if (ibuf->rect == NULL && ibuf->rect_float == NULL) {
+  if (ibuf->byte_buffer.data == NULL && ibuf->float_buffer.data == NULL) {
    return;
  }

@ -1692,7 +1692,7 @@ void BKE_scopes_update(Scopes *scopes,
  scopes->vecscope = MEM_callocN(scopes->waveform_tot * 2 * sizeof(float),
                                 "vectorscope point channel");

-  if (ibuf->rect_float) {
+  if (ibuf->float_buffer.data) {
    cm_processor = IMB_colormanagement_display_processor_new(view_settings, display_settings);
  }
  else {
--- a/source/blender/blenkernel/intern/constraint.c
+++ b/source/blender/blenkernel/intern/constraint.c
@ -6556,7 +6556,7 @@ static void lib_link_constraint_cb(bConstraint *UNUSED(con),
                                   void *userdata)
 {
  tConstraintLinkData *cld = (tConstraintLinkData *)userdata;
-  BLO_read_id_address(cld->reader, cld->id->lib, idpoin);
+  BLO_read_id_address(cld->reader, cld->id, idpoin);
 }

 void BKE_constraint_blend_read_lib(BlendLibReader *reader, ID *id, ListBase *conlist)
@ -6571,7 +6571,7 @@ void BKE_constraint_blend_read_lib(BlendLibReader *reader, ID *id, ListBase *con
      con->type = CONSTRAINT_TYPE_NULL;
    }
    /* own ipo, all constraints have it */
-    BLO_read_id_address(reader, id->lib, &con->ipo); /* XXX deprecated - old animation system */
+    BLO_read_id_address(reader, id, &con->ipo); /* XXX deprecated - old animation system */

    /* If linking from a library, clear 'local' library override flag. */
    if (ID_IS_LINKED(id)) {
--- a/source/blender/blenkernel/intern/curve.cc
+++ b/source/blender/blenkernel/intern/curve.cc
@ -270,19 +270,19 @@ static void curve_blend_read_lib(BlendLibReader *reader, ID *id)
 {
  Curve *cu = (Curve *)id;
  for (int a = 0; a < cu->totcol; a++) {
-    BLO_read_id_address(reader, cu->id.lib, &cu->mat[a]);
+    BLO_read_id_address(reader, id, &cu->mat[a]);
  }

-  BLO_read_id_address(reader, cu->id.lib, &cu->bevobj);
-  BLO_read_id_address(reader, cu->id.lib, &cu->taperobj);
-  BLO_read_id_address(reader, cu->id.lib, &cu->textoncurve);
-  BLO_read_id_address(reader, cu->id.lib, &cu->vfont);
-  BLO_read_id_address(reader, cu->id.lib, &cu->vfontb);
-  BLO_read_id_address(reader, cu->id.lib, &cu->vfonti);
-  BLO_read_id_address(reader, cu->id.lib, &cu->vfontbi);
+  BLO_read_id_address(reader, id, &cu->bevobj);
+  BLO_read_id_address(reader, id, &cu->taperobj);
+  BLO_read_id_address(reader, id, &cu->textoncurve);
+  BLO_read_id_address(reader, id, &cu->vfont);
+  BLO_read_id_address(reader, id, &cu->vfontb);
+  BLO_read_id_address(reader, id, &cu->vfonti);
+  BLO_read_id_address(reader, id, &cu->vfontbi);

-  BLO_read_id_address(reader, cu->id.lib, &cu->ipo); /* XXX deprecated - old animation system */
-  BLO_read_id_address(reader, cu->id.lib, &cu->key);
+  BLO_read_id_address(reader, id, &cu->ipo); /* XXX deprecated - old animation system */
+  BLO_read_id_address(reader, id, &cu->key);
 }

 static void curve_blend_read_expand(BlendExpander *expander, ID *id)
--- a/source/blender/blenkernel/intern/curve_to_mesh_convert.cc
+++ b/source/blender/blenkernel/intern/curve_to_mesh_convert.cc
@ -770,7 +770,7 @@ Mesh *curve_to_mesh_sweep(const CurvesGeometry &main,
    mesh->tag_loose_verts_none();
    if (!offsets.any_single_point_profile) {
      /* If there are no single point profiles, every combination will have faces. */
-      mesh->loose_edges_tag_none();
+      mesh->tag_loose_edges_none();
    }
  }

--- a/source/blender/blenkernel/intern/curves.cc
+++ b/source/blender/blenkernel/intern/curves.cc
@ -137,9 +137,9 @@ static void curves_blend_read_lib(BlendLibReader *reader, ID *id)
 {
  Curves *curves = (Curves *)id;
  for (int a = 0; a < curves->totcol; a++) {
-    BLO_read_id_address(reader, curves->id.lib, &curves->mat[a]);
+    BLO_read_id_address(reader, id, &curves->mat[a]);
  }
-  BLO_read_id_address(reader, curves->id.lib, &curves->surface);
+  BLO_read_id_address(reader, id, &curves->surface);
 }

 static void curves_blend_read_expand(BlendExpander *expander, ID *id)
--- a/source/blender/blenkernel/intern/customdata.cc
+++ b/source/blender/blenkernel/intern/customdata.cc
@ -3176,6 +3176,13 @@ void CustomData_free_layers(CustomData *data, const eCustomDataType type, const
  }
 }

+bool CustomData_has_layer_named(const CustomData *data,
+                                const eCustomDataType type,
+                                const char *name)
+{
+  return CustomData_get_named_layer_index(data, type, name) != -1;
+}
+
 bool CustomData_has_layer(const CustomData *data, const eCustomDataType type)
 {
  return (CustomData_get_layer_index(data, type) != -1);
@ -5181,9 +5188,6 @@ void CustomData_blend_write(BlendWriter *writer,
      case CD_PAINT_MASK:
        BLO_write_raw(writer, sizeof(float) * count, static_cast<const float *>(layer.data));
        break;
-      case CD_SCULPT_FACE_SETS:
-        BLO_write_raw(writer, sizeof(float) * count, static_cast<const float *>(layer.data));
-        break;
      case CD_GRID_PAINT_MASK:
        write_grid_paint_mask(writer, count, static_cast<const GridPaintMask *>(layer.data));
        break;
--- a/source/blender/blenkernel/intern/data_transfer.cc
+++ b/source/blender/blenkernel/intern/data_transfer.cc
@ -1004,7 +1004,7 @@ static bool data_transfer_layersmapping_generate(ListBase *r_map,
      return true;
    }
    if (r_map && cddata_type == CD_FAKE_SEAM) {
-      if (!CustomData_get_layer_named(&me_dst->edata, CD_PROP_BOOL, ".uv_seam")) {
+      if (!CustomData_has_layer_named(&me_dst->edata, CD_PROP_BOOL, ".uv_seam")) {
        CustomData_add_layer_named(
            &me_dst->edata, CD_PROP_BOOL, CD_SET_DEFAULT, me_dst->totedge, ".uv_seam");
      }
@ -1022,7 +1022,7 @@ static bool data_transfer_layersmapping_generate(ListBase *r_map,
      return true;
    }
    if (r_map && cddata_type == CD_FAKE_SHARP) {
-      if (!CustomData_get_layer_named(&me_dst->edata, CD_PROP_BOOL, "sharp_edge")) {
+      if (!CustomData_has_layer_named(&me_dst->edata, CD_PROP_BOOL, "sharp_edge")) {
        CustomData_add_layer_named(
            &me_dst->edata, CD_PROP_BOOL, CD_SET_DEFAULT, me_dst->totedge, "sharp_edge");
      }
@ -1110,7 +1110,7 @@ static bool data_transfer_layersmapping_generate(ListBase *r_map,
      return true;
    }
    if (r_map && cddata_type == CD_FAKE_SHARP) {
-      if (!CustomData_get_layer_named(&me_dst->pdata, CD_PROP_BOOL, "sharp_face")) {
+      if (!CustomData_has_layer_named(&me_dst->pdata, CD_PROP_BOOL, "sharp_face")) {
        CustomData_add_layer_named(
            &me_dst->pdata, CD_PROP_BOOL, CD_SET_DEFAULT, me_dst->totpoly, "sharp_face");
      }
--- a/source/blender/blenkernel/intern/dynamicpaint.cc
+++ b/source/blender/blenkernel/intern/dynamicpaint.cc
@ -3210,12 +3210,15 @@ static void dynamic_paint_output_surface_image_paint_cb(void *__restrict userdat
  const int pos = ((ImgSeqFormatData *)(surface->data->format_data))->uv_p[index].pixel_index * 4;

  /* blend wet and dry layers */
-  blendColors(
-      point->color, point->color[3], point->e_color, point->e_color[3], &ibuf->rect_float[pos]);
+  blendColors(point->color,
+              point->color[3],
+              point->e_color,
+              point->e_color[3],
+              &ibuf->float_buffer.data[pos]);

  /* Multiply color by alpha if enabled */
  if (surface->flags & MOD_DPAINT_MULALPHA) {
-    mul_v3_fl(&ibuf->rect_float[pos], ibuf->rect_float[pos + 3]);
+    mul_v3_fl(&ibuf->float_buffer.data[pos], ibuf->float_buffer.data[pos + 3]);
  }
 }

@ -3242,8 +3245,8 @@ static void dynamic_paint_output_surface_image_displace_cb(

  CLAMP(depth, 0.0f, 1.0f);

-  copy_v3_fl(&ibuf->rect_float[pos], depth);
-  ibuf->rect_float[pos + 3] = 1.0f;
+  copy_v3_fl(&ibuf->float_buffer.data[pos], depth);
+  ibuf->float_buffer.data[pos + 3] = 1.0f;
 }

 static void dynamic_paint_output_surface_image_wave_cb(void *__restrict userdata,
@ -3268,8 +3271,8 @@ static void dynamic_paint_output_surface_image_wave_cb(void *__restrict userdata
  depth = (0.5f + depth / 2.0f);
  CLAMP(depth, 0.0f, 1.0f);

-  copy_v3_fl(&ibuf->rect_float[pos], depth);
-  ibuf->rect_float[pos + 3] = 1.0f;
+  copy_v3_fl(&ibuf->float_buffer.data[pos], depth);
+  ibuf->float_buffer.data[pos + 3] = 1.0f;
 }

 static void dynamic_paint_output_surface_image_wetmap_cb(void *__restrict userdata,
@ -3286,8 +3289,8 @@ static void dynamic_paint_output_surface_image_wetmap_cb(void *__restrict userda
  /* image buffer position */
  const int pos = ((ImgSeqFormatData *)(surface->data->format_data))->uv_p[index].pixel_index * 4;

-  copy_v3_fl(&ibuf->rect_float[pos], (point->wetness > 1.0f) ? 1.0f : point->wetness);
-  ibuf->rect_float[pos + 3] = 1.0f;
+  copy_v3_fl(&ibuf->float_buffer.data[pos], (point->wetness > 1.0f) ? 1.0f : point->wetness);
+  ibuf->float_buffer.data[pos + 3] = 1.0f;
 }

 void dynamicPaint_outputSurfaceImage(DynamicPaintSurface *surface,
--- a/source/blender/blenkernel/intern/fcurve.c
+++ b/source/blender/blenkernel/intern/fcurve.c
@ -2540,7 +2540,7 @@ void BKE_fmodifiers_blend_read_lib(BlendLibReader *reader, ID *id, ListBase *fmo
    switch (fcm->type) {
      case FMODIFIER_TYPE_PYTHON: {
        FMod_Python *data = (FMod_Python *)fcm->data;
-        BLO_read_id_address(reader, id->lib, &data->script);
+        BLO_read_id_address(reader, id, &data->script);
        break;
      }
    }
@ -2670,7 +2670,7 @@ void BKE_fcurve_blend_read_lib(BlendLibReader *reader, ID *id, ListBase *fcurves
        DRIVER_TARGETS_LOOPER_BEGIN (dvar) {
          /* only relink if still used */
          if (tarIndex < dvar->num_targets) {
-            BLO_read_id_address(reader, id->lib, &dtar->id);
+            BLO_read_id_address(reader, id, &dtar->id);
          }
          else {
            dtar->id = NULL;
--- a/source/blender/blenkernel/intern/geometry_set_instances.cc
+++ b/source/blender/blenkernel/intern/geometry_set_instances.cc
@ -3,28 +3,16 @@
 #include "BKE_collection.h"
 #include "BKE_geometry_set_instances.hh"
 #include "BKE_instances.hh"
-#include "BKE_material.h"
 #include "BKE_mesh.hh"
 #include "BKE_mesh_wrapper.h"
 #include "BKE_modifier.h"
-#include "BKE_pointcloud.h"

 #include "DNA_collection_types.h"
 #include "DNA_layer_types.h"
-#include "DNA_mesh_types.h"
-#include "DNA_meshdata_types.h"
 #include "DNA_object_types.h"

 namespace blender::bke {

-static void geometry_set_collect_recursive(const GeometrySet &geometry_set,
-                                           const float4x4 &transform,
-                                           Vector<GeometryInstanceGroup> &r_sets);
-
-static void geometry_set_collect_recursive_collection(const Collection &collection,
-                                                      const float4x4 &transform,
-                                                      Vector<GeometryInstanceGroup> &r_sets);
-
 static void add_final_mesh_as_geometry_component(const Object &object, GeometrySet &geometry_set)
 {
  Mesh *mesh = BKE_modifier_get_evaluated_mesh_from_evaluated_object(
@ -74,87 +62,6 @@ GeometrySet object_get_evaluated_geometry_set(const Object &object)
  return {};
 }

-static void geometry_set_collect_recursive_collection_instance(
-    const Collection &collection, const float4x4 &transform, Vector<GeometryInstanceGroup> &r_sets)
-{
-  float4x4 offset_matrix = float4x4::identity();
-  offset_matrix.location() -= float3(collection.instance_offset);
-  const float4x4 instance_transform = transform * offset_matrix;
-  geometry_set_collect_recursive_collection(collection, instance_transform, r_sets);
-}
-
-static void geometry_set_collect_recursive_object(const Object &object,
-                                                  const float4x4 &transform,
-                                                  Vector<GeometryInstanceGroup> &r_sets)
-{
-  GeometrySet instance_geometry_set = object_get_evaluated_geometry_set(object);
-  geometry_set_collect_recursive(instance_geometry_set, transform, r_sets);
-}
-
-static void geometry_set_collect_recursive_collection(const Collection &collection,
-                                                      const float4x4 &transform,
-                                                      Vector<GeometryInstanceGroup> &r_sets)
-{
-  LISTBASE_FOREACH (const CollectionObject *, collection_object, &collection.gobject) {
-    BLI_assert(collection_object->ob != nullptr);
-    const Object &object = *collection_object->ob;
-    const float4x4 object_transform = transform * float4x4_view(object.object_to_world);
-    geometry_set_collect_recursive_object(object, object_transform, r_sets);
-  }
-  LISTBASE_FOREACH (const CollectionChild *, collection_child, &collection.children) {
-    BLI_assert(collection_child->collection != nullptr);
-    const Collection &collection = *collection_child->collection;
-    geometry_set_collect_recursive_collection(collection, transform, r_sets);
-  }
-}
-
-static void geometry_set_collect_recursive(const GeometrySet &geometry_set,
-                                           const float4x4 &transform,
-                                           Vector<GeometryInstanceGroup> &r_sets)
-{
-  r_sets.append({geometry_set, {transform}});
-
-  if (geometry_set.has_instances()) {
-    const Instances &instances = *geometry_set.get_instances_for_read();
-
-    Span<float4x4> transforms = instances.transforms();
-    Span<int> handles = instances.reference_handles();
-    Span<InstanceReference> references = instances.references();
-    for (const int i : transforms.index_range()) {
-      const InstanceReference &reference = references[handles[i]];
-      const float4x4 instance_transform = transform * transforms[i];
-
-      switch (reference.type()) {
-        case InstanceReference::Type::Object: {
-          Object &object = reference.object();
-          geometry_set_collect_recursive_object(object, instance_transform, r_sets);
-          break;
-        }
-        case InstanceReference::Type::Collection: {
-          Collection &collection = reference.collection();
-          geometry_set_collect_recursive_collection_instance(
-              collection, instance_transform, r_sets);
-          break;
-        }
-        case InstanceReference::Type::GeometrySet: {
-          const GeometrySet &geometry_set = reference.geometry_set();
-          geometry_set_collect_recursive(geometry_set, instance_transform, r_sets);
-          break;
-        }
-        case InstanceReference::Type::None: {
-          break;
-        }
-      }
-    }
-  }
-}
-
-void geometry_set_gather_instances(const GeometrySet &geometry_set,
-                                   Vector<GeometryInstanceGroup> &r_instance_groups)
-{
-  geometry_set_collect_recursive(geometry_set, float4x4::identity(), r_instance_groups);
-}
-
 void Instances::foreach_referenced_geometry(
    blender::FunctionRef<void(const GeometrySet &geometry_set)> callback) const
 {
--- a/source/blender/blenkernel/intern/gpencil_legacy.c
+++ b/source/blender/blenkernel/intern/gpencil_legacy.c
@ -275,12 +275,12 @@ static void greasepencil_blend_read_lib(BlendLibReader *reader, ID *id)
  /* Layers */
  LISTBASE_FOREACH (bGPDlayer *, gpl, &gpd->layers) {
    /* Layer -> Parent References */
-    BLO_read_id_address(reader, gpd->id.lib, &gpl->parent);
+    BLO_read_id_address(reader, id, &gpl->parent);
  }

  /* materials */
  for (int a = 0; a < gpd->totcol; a++) {
-    BLO_read_id_address(reader, gpd->id.lib, &gpd->mat[a]);
+    BLO_read_id_address(reader, id, &gpd->mat[a]);
  }
 }

@ -2221,8 +2221,7 @@ int BKE_gpencil_object_material_index_get_by_name(Object *ob, const char *name)
  Material *read_ma = NULL;
  for (short i = 0; i < *totcol; i++) {
    read_ma = BKE_object_material_get(ob, i + 1);
-    /* Material names are like "MAMaterial.001" */
-    if (STREQ(name, &read_ma->id.name[2])) {
+    if (STREQ(name, read_ma->id.name + 2)) {
      return i;
    }
  }
@ -2308,7 +2307,7 @@ bool BKE_gpencil_from_image(

  ibuf = BKE_image_acquire_ibuf(image, &iuser, &lock);

-  if (ibuf && ibuf->rect) {
+  if (ibuf && ibuf->byte_buffer.data) {
    int img_x = ibuf->x;
    int img_y = ibuf->y;

--- a/source/blender/blenkernel/intern/icons.cc
+++ b/source/blender/blenkernel/intern/icons.cc
@ -552,7 +552,7 @@ void BKE_previewimg_ensure(PreviewImage *prv, const int size)
  if (do_preview) {
    prv->w[ICON_SIZE_PREVIEW] = thumb->x;
    prv->h[ICON_SIZE_PREVIEW] = thumb->y;
-    prv->rect[ICON_SIZE_PREVIEW] = (uint *)MEM_dupallocN(thumb->rect);
+    prv->rect[ICON_SIZE_PREVIEW] = (uint *)MEM_dupallocN(thumb->byte_buffer.data);
    prv->flag[ICON_SIZE_PREVIEW] &= ~(PRV_CHANGED | PRV_USER_EDITED | PRV_RENDERING);
  }
  if (do_icon) {
@ -571,7 +571,7 @@ void BKE_previewimg_ensure(PreviewImage *prv, const int size)
    IMB_scaleImBuf(thumb, icon_w, icon_h);
    prv->w[ICON_SIZE_ICON] = icon_w;
    prv->h[ICON_SIZE_ICON] = icon_h;
-    prv->rect[ICON_SIZE_ICON] = (uint *)MEM_dupallocN(thumb->rect);
+    prv->rect[ICON_SIZE_ICON] = (uint *)MEM_dupallocN(thumb->byte_buffer.data);
    prv->flag[ICON_SIZE_ICON] &= ~(PRV_CHANGED | PRV_USER_EDITED | PRV_RENDERING);
  }
  IMB_freeImBuf(thumb);
@ -588,7 +588,7 @@ ImBuf *BKE_previewimg_to_imbuf(PreviewImage *prv, const int size)
  if (w > 0 && h > 0 && rect) {
    /* first allocate imbuf for copying preview into it */
    ima = IMB_allocImBuf(w, h, 32, IB_rect);
-    memcpy(ima->rect, rect, w * h * sizeof(*ima->rect));
+    memcpy(ima->byte_buffer.data, rect, w * h * sizeof(uint8_t) * 4);
  }

  return ima;
--- a/source/blender/blenkernel/intern/icons_rasterize.c
+++ b/source/blender/blenkernel/intern/icons_rasterize.c
@ -85,7 +85,7 @@ ImBuf *BKE_icon_geom_rasterize(const struct Icon_Geom *geom, const uint size_x,
  data.rect_size[0] = rect_size[0];
  data.rect_size[1] = rect_size[1];

-  data.rect = ibuf->rect;
+  data.rect = (uint *)ibuf->byte_buffer.data;

  float scale[2];
  const bool use_scale = (rect_size[0] != 256) || (rect_size[1] != 256);
--- a/source/blender/blenkernel/intern/idprop.c
+++ b/source/blender/blenkernel/intern/idprop.c
@ -1495,7 +1495,7 @@ void IDP_BlendReadData_impl(BlendDataReader *reader, IDProperty **prop, const ch
  }
 }

-void IDP_BlendReadLib(BlendLibReader *reader, Library *lib, IDProperty *prop)
+void IDP_BlendReadLib(BlendLibReader *reader, ID *self_id, IDProperty *prop)
 {
  if (!prop) {
    return;
@ -1504,7 +1504,8 @@ void IDP_BlendReadLib(BlendLibReader *reader, Library *lib, IDProperty *prop)
  switch (prop->type) {
    case IDP_ID: /* PointerProperty */
    {
-      void *newaddr = BLO_read_get_new_id_address(reader, lib, IDP_Id(prop));
+      void *newaddr = BLO_read_get_new_id_address(
+          reader, self_id, ID_IS_LINKED(self_id), IDP_Id(prop));
      if (IDP_Id(prop) && !newaddr && G.debug) {
        printf("Error while loading \"%s\". Data not found in file!\n", prop->name);
      }
@ -1515,14 +1516,14 @@ void IDP_BlendReadLib(BlendLibReader *reader, Library *lib, IDProperty *prop)
    {
      IDProperty *idp_array = IDP_IDPArray(prop);
      for (int i = 0; i < prop->len; i++) {
-        IDP_BlendReadLib(reader, lib, &(idp_array[i]));
+        IDP_BlendReadLib(reader, self_id, &(idp_array[i]));
      }
      break;
    }
    case IDP_GROUP: /* PointerProperty */
    {
      LISTBASE_FOREACH (IDProperty *, loop, &prop->data.group) {
-        IDP_BlendReadLib(reader, lib, loop);
+        IDP_BlendReadLib(reader, self_id, loop);
      }
      break;
    }
--- a/source/blender/blenkernel/intern/image.cc
+++ b/source/blender/blenkernel/intern/image.cc
@ -1138,7 +1138,7 @@ static ImBuf *add_ibuf_for_tile(Image *ima, ImageTile *tile)
    }

    if (ibuf != nullptr) {
-      rect_float = ibuf->rect_float;
+      rect_float = ibuf->float_buffer.data;
      IMB_colormanagement_check_is_data(ibuf, ima->colorspace_settings.name);
    }

@ -1162,7 +1162,7 @@ static ImBuf *add_ibuf_for_tile(Image *ima, ImageTile *tile)
    }

    if (ibuf != nullptr) {
-      rect = (uchar *)ibuf->rect;
+      rect = ibuf->byte_buffer.data;
      IMB_colormanagement_assign_rect_colorspace(ibuf, ima->colorspace_settings.name);
    }

@ -1261,7 +1261,7 @@ static void image_colorspace_from_imbuf(Image *image, const ImBuf *ibuf)
 {
  const char *colorspace_name = nullptr;

-  if (ibuf->rect_float) {
+  if (ibuf->float_buffer.data) {
    if (ibuf->float_colorspace) {
      colorspace_name = IMB_colormanagement_colorspace_get_name(ibuf->float_colorspace);
    }
@ -1270,7 +1270,7 @@ static void image_colorspace_from_imbuf(Image *image, const ImBuf *ibuf)
    }
  }

-  if (ibuf->rect && !colorspace_name) {
+  if (ibuf->byte_buffer.data && !colorspace_name) {
    if (ibuf->rect_colorspace) {
      colorspace_name = IMB_colormanagement_colorspace_get_name(ibuf->rect_colorspace);
    }
@ -1318,7 +1318,7 @@ void BKE_image_replace_imbuf(Image *image, ImBuf *ibuf)

  /* Keep generated image type flags consistent with the image buffer. */
  if (image->source == IMA_SRC_GENERATED) {
-    if (ibuf->rect_float) {
+    if (ibuf->float_buffer.data) {
      image->gen_flag |= IMA_GEN_FLOAT;
    }
    else {
@ -1338,11 +1338,11 @@ void BKE_image_replace_imbuf(Image *image, ImBuf *ibuf)
 static bool image_memorypack_imbuf(
    Image *ima, ImBuf *ibuf, int view, int tile_number, const char *filepath)
 {
-  ibuf->ftype = (ibuf->rect_float) ? IMB_FTYPE_OPENEXR : IMB_FTYPE_PNG;
+  ibuf->ftype = (ibuf->float_buffer.data) ? IMB_FTYPE_OPENEXR : IMB_FTYPE_PNG;

  IMB_saveiff(ibuf, filepath, IB_rect | IB_mem);

-  if (ibuf->encodedbuffer == nullptr) {
+  if (ibuf->encoded_buffer.data == nullptr) {
    CLOG_STR_ERROR(&LOG, "memory save for pack error");
    IMB_freeImBuf(ibuf);
    image_free_packedfiles(ima);
@ -1352,8 +1352,8 @@ static bool image_memorypack_imbuf(
  ImagePackedFile *imapf;
  PackedFile *pf = MEM_cnew<PackedFile>("PackedFile");

-  pf->data = ibuf->encodedbuffer;
-  pf->size = ibuf->encodedsize;
+  pf->data = IMB_steal_encoded_buffer(ibuf);
+  pf->size = ibuf->encoded_size;

  imapf = static_cast<ImagePackedFile *>(MEM_mallocN(sizeof(ImagePackedFile), "Image PackedFile"));
  STRNCPY(imapf->filepath, filepath);
@ -1362,8 +1362,6 @@ static bool image_memorypack_imbuf(
  imapf->tile_number = tile_number;
  BLI_addtail(&ima->packedfiles, imapf);

-  ibuf->encodedbuffer = nullptr;
-  ibuf->encodedsize = 0;
  ibuf->userflags &= ~IB_BITMAPDIRTY;

  return true;
@ -1505,26 +1503,12 @@ static uintptr_t image_mem_size(Image *image)
      if (ibuf == nullptr) {
        continue;
      }
-      ImBuf *ibufm;
-      int level;

-      if (ibuf->rect) {
-        size += MEM_allocN_len(ibuf->rect);
-      }
-      if (ibuf->rect_float) {
-        size += MEM_allocN_len(ibuf->rect_float);
-      }
+      size += IMB_get_size_in_memory(ibuf);

-      for (level = 0; level < IMB_MIPMAP_LEVELS; level++) {
-        ibufm = ibuf->mipmap[level];
-        if (ibufm) {
-          if (ibufm->rect) {
-            size += MEM_allocN_len(ibufm->rect);
-          }
-          if (ibufm->rect_float) {
-            size += MEM_allocN_len(ibufm->rect_float);
-          }
-        }
+      for (int level = 0; level < IMB_MIPMAP_LEVELS; level++) {
+        ImBuf *ibufm = ibuf->mipmap[level];
+        size += IMB_get_size_in_memory(ibufm);
      }
    }
    IMB_moviecacheIter_free(iter);
@ -2537,16 +2521,16 @@ void BKE_stamp_info_from_imbuf(RenderResult *rr, ImBuf *ibuf)
 bool BKE_imbuf_alpha_test(ImBuf *ibuf)
 {
  int tot;
-  if (ibuf->rect_float) {
-    const float *buf = ibuf->rect_float;
+  if (ibuf->float_buffer.data) {
+    const float *buf = ibuf->float_buffer.data;
    for (tot = ibuf->x * ibuf->y; tot--; buf += 4) {
      if (buf[3] < 1.0f) {
        return true;
      }
    }
  }
-  else if (ibuf->rect) {
-    uchar *buf = (uchar *)ibuf->rect;
+  else if (ibuf->byte_buffer.data) {
+    uchar *buf = ibuf->byte_buffer.data;
    for (tot = ibuf->x * ibuf->y; tot--; buf += 4) {
      if (buf[3] != 255) {
        return true;
@ -4007,9 +3991,8 @@ static ImBuf *image_load_sequence_multilayer(Image *ima, ImageUser *iuser, int e
      // printf("load from pass %s\n", rpass->name);
      /* since we free  render results, we copy the rect */
      ibuf = IMB_allocImBuf(ima->rr->rectx, ima->rr->recty, 32, 0);
-      ibuf->rect_float = static_cast<float *>(MEM_dupallocN(rpass->rect));
-      ibuf->flags |= IB_rectfloat;
-      ibuf->mall = IB_rectfloat;
+      IMB_assign_float_buffer(
+          ibuf, static_cast<float *>(MEM_dupallocN(rpass->rect)), IB_TAKE_OWNERSHIP);
      ibuf->channels = rpass->channels;

      BKE_imbuf_stamp_info(ima->rr, ibuf);
@ -4318,8 +4301,7 @@ static ImBuf *image_get_ibuf_multilayer(Image *ima, ImageUser *iuser)

      image_init_after_load(ima, iuser, ibuf);

-      ibuf->rect_float = rpass->rect;
-      ibuf->flags |= IB_rectfloat;
+      IMB_assign_float_buffer(ibuf, rpass->rect, IB_DO_NOT_TAKE_OWNERSHIP);
      ibuf->channels = rpass->channels;

      BKE_imbuf_stamp_info(ima->rr, ibuf);
@ -4459,56 +4441,29 @@ static ImBuf *image_get_render_result(Image *ima, ImageUser *iuser, void **r_loc
   *
   * For other cases we need to be sure it stays to default byte buffer space.
   */
-  if (ibuf->rect != rect) {
+  if (ibuf->byte_buffer.data != (uint8_t *)rect) {
    const char *colorspace = IMB_colormanagement_role_colorspace_name_get(COLOR_ROLE_DEFAULT_BYTE);
    IMB_colormanagement_assign_rect_colorspace(ibuf, colorspace);
  }

  /* invalidate color managed buffers if render result changed */
  BLI_thread_lock(LOCK_COLORMANAGE);
-  if (ibuf->x != rres.rectx || ibuf->y != rres.recty || ibuf->rect_float != rectf) {
+  if (ibuf->x != rres.rectx || ibuf->y != rres.recty || ibuf->float_buffer.data != rectf) {
    ibuf->userflags |= IB_DISPLAY_BUFFER_INVALID;
  }

  ibuf->x = rres.rectx;
  ibuf->y = rres.recty;
+  ibuf->channels = channels;

-  if (rect) {
-    imb_freerectImBuf(ibuf);
-    ibuf->rect = rect;
-  }
-  else {
-    /* byte buffer of render result has been freed, make sure image buffers
-     * does not reference to this buffer anymore
-     * need check for whether byte buffer was allocated and owned by image itself
-     * or if it's reusing buffer from render result
-     */
-    if ((ibuf->mall & IB_rect) == 0) {
-      ibuf->rect = nullptr;
-    }
-  }
+  imb_freerectImBuf(ibuf);

-  if (rectf) {
-    ibuf->rect_float = rectf;
-    ibuf->flags |= IB_rectfloat;
-    ibuf->channels = channels;
-  }
-  else {
-    ibuf->rect_float = nullptr;
-    ibuf->flags &= ~IB_rectfloat;
-  }
-
-  if (rectz) {
-    ibuf->zbuf_float = rectz;
-    ibuf->flags |= IB_zbuffloat;
-  }
-  else {
-    ibuf->zbuf_float = nullptr;
-    ibuf->flags &= ~IB_zbuffloat;
-  }
+  IMB_assign_byte_buffer(ibuf, (uint8_t *)rect, IB_DO_NOT_TAKE_OWNERSHIP);
+  IMB_assign_float_buffer(ibuf, rectf, IB_DO_NOT_TAKE_OWNERSHIP);
+  IMB_assign_float_z_buffer(ibuf, rectz, IB_DO_NOT_TAKE_OWNERSHIP);

  /* TODO(sergey): Make this faster by either simply referencing the stamp
-   * or by changing both ImBug and RenderResult to use same data type to
+   * or by changing both ImBuf and RenderResult to use same data type to
   * store metadata. */
  if (ibuf->metadata != nullptr) {
    IMB_metadata_free(ibuf->metadata);
@ -5269,7 +5224,7 @@ uchar *BKE_image_get_pixels_for_frame(struct Image *image, int frame, int tile)
  ibuf = BKE_image_acquire_ibuf(image, &iuser, &lock);

  if (ibuf) {
-    pixels = (uchar *)ibuf->rect;
+    pixels = ibuf->byte_buffer.data;

    if (pixels) {
      pixels = static_cast<uchar *>(MEM_dupallocN(pixels));
@ -5299,7 +5254,7 @@ float *BKE_image_get_float_pixels_for_frame(struct Image *image, int frame, int
  ibuf = BKE_image_acquire_ibuf(image, &iuser, &lock);

  if (ibuf) {
-    pixels = ibuf->rect_float;
+    pixels = ibuf->float_buffer.data;

    if (pixels) {
      pixels = static_cast<float *>(MEM_dupallocN(pixels));
--- a/source/blender/blenkernel/intern/image_format.cc
+++ b/source/blender/blenkernel/intern/image_format.cc
@ -799,7 +799,7 @@ void BKE_image_format_from_imbuf(ImageFormatData *im_format, const ImBuf *imbuf)
    if (custom_flags & OPENEXR_COMPRESS) {
      im_format->exr_codec = R_IMF_EXR_CODEC_ZIP; /* Can't determine compression */
    }
-    if (imbuf->zbuf_float) {
+    if (imbuf->float_z_buffer.data) {
      im_format->flag |= R_IMF_FLAG_ZBUF;
    }
  }
--- a/source/blender/blenkernel/intern/image_gpu.cc
+++ b/source/blender/blenkernel/intern/image_gpu.cc
@ -50,11 +50,11 @@ bool BKE_image_has_gpu_texture_premultiplied_alpha(Image *image, ImBuf *ibuf)
    }
    /* Generated images use pre multiplied float buffer, but straight alpha for byte buffers. */
    if (image->type == IMA_TYPE_UV_TEST && ibuf) {
-      return ibuf->rect_float != nullptr;
+      return ibuf->float_buffer.data != nullptr;
    }
  }
  if (ibuf) {
-    if (ibuf->rect_float) {
+    if (ibuf->float_buffer.data) {
      return image ? (image->alpha_mode != IMA_ALPHA_STRAIGHT) : false;
    }

@ -643,7 +643,7 @@ static ImBuf *update_do_scale(uchar *rect,
  }

  /* Scale pixels. */
-  ImBuf *ibuf = IMB_allocFromBuffer((uint *)rect, rect_float, part_w, part_h, 4);
+  ImBuf *ibuf = IMB_allocFromBuffer(rect, rect_float, part_w, part_h, 4);
  IMB_scaleImBuf(ibuf, *w, *h);

  return ibuf;
@ -679,8 +679,9 @@ static void gpu_texture_update_scaled(GPUTexture *tex,
    ibuf = update_do_scale(rect, rect_float, &x, &y, &w, &h, limit_w, limit_h, full_w, full_h);
  }

-  void *data = (ibuf->rect_float) ? (void *)(ibuf->rect_float) : (void *)(ibuf->rect);
-  eGPUDataFormat data_format = (ibuf->rect_float) ? GPU_DATA_FLOAT : GPU_DATA_UBYTE;
+  void *data = (ibuf->float_buffer.data) ? (void *)(ibuf->float_buffer.data) :
+                                           (void *)(ibuf->byte_buffer.data);
+  eGPUDataFormat data_format = (ibuf->float_buffer.data) ? GPU_DATA_FLOAT : GPU_DATA_UBYTE;

  GPU_texture_update_sub(tex, data_format, data, x, y, layer, w, h, 1);

@ -742,8 +743,8 @@ static void gpu_texture_update_from_ibuf(
  }

  /* Get texture data pointers. */
-  float *rect_float = ibuf->rect_float;
-  uchar *rect = (uchar *)ibuf->rect;
+  float *rect_float = ibuf->float_buffer.data;
+  uchar *rect = ibuf->byte_buffer.data;
  int tex_stride = ibuf->x;
  int tex_offset = ibuf->channels * (y * ibuf->x + x);

@ -832,10 +833,10 @@ static void gpu_texture_update_from_ibuf(
  }

  /* Free buffers if needed. */
-  if (rect && rect != (uchar *)ibuf->rect) {
+  if (rect && rect != ibuf->byte_buffer.data) {
    MEM_freeN(rect);
  }
-  if (rect_float && rect_float != ibuf->rect_float) {
+  if (rect_float && rect_float != ibuf->float_buffer.data) {
    MEM_freeN(rect_float);
  }

--- a/source/blender/blenkernel/intern/image_save.cc
+++ b/source/blender/blenkernel/intern/image_save.cc
@ -42,7 +42,7 @@ static char imtype_best_depth(ImBuf *ibuf, const char imtype)
 {
  const char depth_ok = BKE_imtype_valid_depths(imtype);

-  if (ibuf->rect_float) {
+  if (ibuf->float_buffer.data) {
    if (depth_ok & R_IMF_CHAN_DEPTH_32) {
      return R_IMF_CHAN_DEPTH_32;
    }
@ -298,18 +298,10 @@ static void image_save_post(ReportList *reports,
    /* workaround to ensure the render result buffer is no longer used
     * by this image, otherwise can crash when a new render result is
     * created. */
-    if (ibuf->rect && !(ibuf->mall & IB_rect)) {
-      imb_freerectImBuf(ibuf);
-    }
-    if (ibuf->rect_float && !(ibuf->mall & IB_rectfloat)) {
-      imb_freerectfloatImBuf(ibuf);
-    }
-    if (ibuf->zbuf && !(ibuf->mall & IB_zbuf)) {
-      IMB_freezbufImBuf(ibuf);
-    }
-    if (ibuf->zbuf_float && !(ibuf->mall & IB_zbuffloat)) {
-      IMB_freezbuffloatImBuf(ibuf);
-    }
+    imb_freerectImBuf(ibuf);
+    imb_freerectfloatImBuf(ibuf);
+    IMB_freezbufImBuf(ibuf);
+    IMB_freezbuffloatImBuf(ibuf);
  }
  if (ELEM(ima->source, IMA_SRC_GENERATED, IMA_SRC_VIEWER)) {
    ima->source = IMA_SRC_FILE;
@ -365,7 +357,8 @@ static bool image_save_single(ReportList *reports,
  RenderResult *rr = nullptr;
  bool ok = false;

-  if (ibuf == nullptr || (ibuf->rect == nullptr && ibuf->rect_float == nullptr)) {
+  if (ibuf == nullptr || (ibuf->byte_buffer.data == nullptr && ibuf->float_buffer.data == nullptr))
+  {
    BKE_image_release_ibuf(ima, ibuf, lock);
    return ok;
  }
--- a/source/blender/blenkernel/intern/ipo.c
+++ b/source/blender/blenkernel/intern/ipo.c
@ -144,7 +144,7 @@ static void ipo_blend_read_lib(BlendLibReader *reader, ID *id)

  LISTBASE_FOREACH (IpoCurve *, icu, &ipo->curve) {
    if (icu->driver) {
-      BLO_read_id_address(reader, ipo->id.lib, &icu->driver->ob);
+      BLO_read_id_address(reader, id, &icu->driver->ob);
    }
  }
 }
--- a/source/blender/blenkernel/intern/key.cc
+++ b/source/blender/blenkernel/intern/key.cc
@ -185,8 +185,8 @@ static void shapekey_blend_read_lib(BlendLibReader *reader, ID *id)
  Key *key = (Key *)id;
  BLI_assert((key->id.tag & LIB_TAG_EXTERN) == 0);

-  BLO_read_id_address(reader, key->id.lib, &key->ipo); /* XXX deprecated - old animation system */
-  BLO_read_id_address(reader, key->id.lib, &key->from);
+  BLO_read_id_address(reader, id, &key->ipo); /* XXX deprecated - old animation system */
+  BLO_read_id_address(reader, id, &key->from);
 }

 static void shapekey_blend_read_expand(BlendExpander *expander, ID *id)
--- a/Show More
+++ b/Show More