Group passes by material

This reverts commit 45bc00d298.

.clang-format

@@ -266,7 +266,6 @@ ForEachMacros:
   - SET_SLOT_PROBING_BEGIN
   - MAP_SLOT_PROBING_BEGIN
   - VECTOR_SET_SLOT_PROBING_BEGIN
-  - TGSET_ITER
   - WL_ARRAY_FOR_EACH
   - FOREACH_SPECTRUM_CHANNEL
 

.gitea/default_merge_message/REBASE_TEMPLATE.md

@@ -1,5 +0,0 @@
-${CommitTitle}
-
-${CommitBody}
-
-Pull Request #${PullRequestIndex}

.gitea/default_merge_message/SQUASH_TEMPLATE.md

@@ -1,3 +0,0 @@
-${PullRequestTitle}
-
-Pull Request #${PullRequestIndex}

.gitea/issue_template/bug.yaml

@@ -1,15 +1,13 @@
 name: Bug Report
 about: File a bug report
 labels:
-  - "type::Report"
+  - bug
-  - "status::Needs Triage"
-  - "priority::Normal"
 body:
   - type: markdown
     attributes:
       value: |
                 ### Instructions
-                First time reporting? See [tips](https://wiki.blender.org/wiki/Process/Bug_Reports).
+                First time reporting? See [tips](https://wiki.blender.org/wiki/Process/Bug_Reports) and [walkthrough video](https://www.youtube.com/watch?v=JTD0OJq_rF4).
 
                 * Use **Help > Report a Bug** in Blender to fill system information and exact Blender version.
                 * Test [daily builds](https://builder.blender.org/) to verify if the issue is already fixed.
@@ -21,7 +19,6 @@ body:
     id: body
     attributes:
       label: "Description"
-      hide_label: true
       value: |
                **System Information**
                Operating system:

.gitea/issue_template/design.yaml

@@ -1,10 +1,9 @@
 name: Design
 about: Create a design task (for developers only)
 labels:
-  - "type::Design"
+  - design
 body:
   - type: textarea
     id: body
     attributes:
       label: "Description"
-      hide_label: true

.gitea/issue_template/todo.yaml

@@ -1,10 +1,9 @@
 name: To Do
 about: Create a to do task (for developers only)
 labels:
-  - "type::To Do"
+  - todo
 body:
   - type: textarea
     id: body
     attributes:
       label: "Description"
-      hide_label: true

.gitea/pull_request_template.yaml

@@ -14,4 +14,7 @@ body:
     id: body
     attributes:
       label: "Description"
-      hide_label: true
+      value: |
+               Description of the problem that is addressed in the patch.
+
+               Description of the proposed solution and its implementation.

CMakeLists.txt

@@ -167,26 +167,14 @@ get_blender_version()
 option(WITH_BLENDER "Build blender (disable to build only the blender player)" ON)
 mark_as_advanced(WITH_BLENDER)
 
-if(WIN32)
+if(APPLE)
-  option(WITH_BLENDER_THUMBNAILER "\
+  # In future, can be used with `quicklookthumbnailing/qlthumbnailreply` to create file
-Build \"BlendThumb.dll\" helper for Windows explorer integration to support extracting \
+  # thumbnails for say Finder. Turn it off for now.
-thumbnails from `.blend` files."
+  option(WITH_BLENDER_THUMBNAILER "Build \"blender-thumbnailer\" thumbnail extraction utility" OFF)
-    ON
+elseif(WIN32)
-  )
+  option(WITH_BLENDER_THUMBNAILER "Build \"BlendThumb.dll\" helper for Windows explorer integration" ON)
 else()
-  set(_option_default ON)
+  option(WITH_BLENDER_THUMBNAILER "Build \"blender-thumbnailer\" thumbnail extraction utility" ON)
-  if(APPLE)
-    # In future, can be used with `quicklookthumbnailing/qlthumbnailreply`
-    # to create file thumbnails for say Finder.
-    # Turn it off for now, even though it can build on APPLE, it's not likely to be useful.
-    set(_option_default OFF)
-  endif()
-  option(WITH_BLENDER_THUMBNAILER "\
-Build stand-alone \"blender-thumbnailer\" command-line thumbnail extraction utility, \
-intended for use by file-managers to extract PNG images from `.blend` files."
-    ${_option_default}
-  )
-  unset(_option_default)
 endif()
 
 option(WITH_INTERNATIONAL "Enable I18N (International fonts and text)" ON)
@@ -226,19 +214,14 @@ option(WITH_BULLET        "Enable Bullet (Physics Engine)" ON)
 option(WITH_SYSTEM_BULLET "Use the systems bullet library (currently unsupported due to missing features in upstream!)" )
 mark_as_advanced(WITH_SYSTEM_BULLET)
 option(WITH_OPENCOLORIO   "Enable OpenColorIO color management" ON)
-
-set(_option_default ON)
 if(APPLE)
   # There's no OpenXR runtime in sight for macOS, neither is code well
   # tested there -> disable it by default.
-  set(_option_default OFF)
+  option(WITH_XR_OPENXR   "Enable VR features through the OpenXR specification" OFF)
-endif()
-option(WITH_XR_OPENXR "Enable VR features through the OpenXR specification" ${_option_default})
-if(APPLE)
   mark_as_advanced(WITH_XR_OPENXR)
+else()
+  option(WITH_XR_OPENXR   "Enable VR features through the OpenXR specification" ON)
 endif()
-unset(_option_default)
-
 option(WITH_GMP "Enable features depending on GMP (Exact Boolean)" ON)
 
 # Compositor
@@ -370,12 +353,11 @@ else()
   set(WITH_COREAUDIO OFF)
 endif()
 if(NOT WIN32)
-  set(_option_default ON)
   if(APPLE)
-    set(_option_default OFF)
+    option(WITH_JACK          "Enable JACK Support (http://www.jackaudio.org)" OFF)
+  else()
+    option(WITH_JACK          "Enable JACK Support (http://www.jackaudio.org)" ON)
   endif()
-  option(WITH_JACK "Enable JACK Support (http://www.jackaudio.org)" ${_option_default})
-  unset(_option_default)
   option(WITH_JACK_DYNLOAD  "Enable runtime dynamic JACK libraries loading" OFF)
 else()
   set(WITH_JACK OFF)
@@ -524,7 +506,7 @@ endif()
 if(NOT APPLE)
   option(WITH_CYCLES_DEVICE_HIP        "Enable Cycles AMD HIP support" ON)
   option(WITH_CYCLES_HIP_BINARIES      "Build Cycles AMD HIP binaries" OFF)
-  set(CYCLES_HIP_BINARIES_ARCH gfx1010 gfx1011 gfx1012 gfx1030 gfx1031 gfx1032 gfx1034 gfx1035 gfx1100 gfx1101 gfx1102 CACHE STRING "AMD HIP architectures to build binaries for")
+  set(CYCLES_HIP_BINARIES_ARCH gfx900 gfx906 gfx90c gfx902 gfx1010 gfx1011 gfx1012 gfx1030 gfx1031 gfx1032 gfx1034 gfx1035 gfx1100 gfx1101 gfx1102 CACHE STRING "AMD HIP architectures to build binaries for")
   mark_as_advanced(WITH_CYCLES_DEVICE_HIP)
   mark_as_advanced(CYCLES_HIP_BINARIES_ARCH)
 endif()
@@ -707,12 +689,6 @@ if(CMAKE_COMPILER_IS_GNUCC OR CMAKE_C_COMPILER_ID MATCHES "Clang")
         [HKEY_LOCAL_MACHINE\\SOFTWARE\\Wow6432Node\\LLVM\\LLVM;]/lib/clang/7.0.0/lib/windows
         [HKEY_LOCAL_MACHINE\\SOFTWARE\\Wow6432Node\\LLVM\\LLVM;]/lib/clang/6.0.0/lib/windows
       )
-        find_library(
-          COMPILER_ASAN_LIBRARY_THUNK NAMES clang_rt.asan_dll_thunk-x86_64
-        PATHS
-        [HKEY_LOCAL_MACHINE\\SOFTWARE\\Wow6432Node\\LLVM\\LLVM;]/lib/clang/7.0.0/lib/windows
-        [HKEY_LOCAL_MACHINE\\SOFTWARE\\Wow6432Node\\LLVM\\LLVM;]/lib/clang/6.0.0/lib/windows
-      )
     elseif(APPLE)
       execute_process(COMMAND ${CMAKE_CXX_COMPILER}
         -print-file-name=lib
@@ -733,7 +709,6 @@ if(CMAKE_COMPILER_IS_GNUCC OR CMAKE_C_COMPILER_ID MATCHES "Clang")
       )
     endif()
 
-    mark_as_advanced(COMPILER_ASAN_LIBRARY_THUNK)
     mark_as_advanced(COMPILER_ASAN_LIBRARY)
   endif()
 endif()
@@ -791,8 +766,6 @@ if("${CMAKE_GENERATOR}" MATCHES "Ninja")
   mark_as_advanced(WITH_NINJA_POOL_JOBS)
 endif()
 
-option(WITH_INSTANT_MESHES, "Instant Meshes Quadrangulator" ON)
-
 # Installation process.
 set(POSTINSTALL_SCRIPT "" CACHE FILEPATH "Run given CMake script after installation process")
 mark_as_advanced(POSTINSTALL_SCRIPT)
@@ -1063,9 +1036,9 @@ if(NOT CMAKE_BUILD_TYPE MATCHES "Release")
       unset(_list_COMPILER_ASAN_CFLAGS)
       unset(_is_CONFIG_DEBUG)
     elseif(COMPILER_ASAN_LIBRARY)
-      set(PLATFORM_LINKLIBS "${PLATFORM_LINKLIBS};\"${COMPILER_ASAN_LIBRARY}\" \"${COMPILER_ASAN_LIBRARY_THUNK}\"")
+      set(PLATFORM_LINKLIBS "${PLATFORM_LINKLIBS};${COMPILER_ASAN_LIBRARY}")
-      set(PLATFORM_LINKFLAGS "\"${COMPILER_ASAN_LIBRARY}\" \"${COMPILER_ASAN_LIBRARY_THUNK}\" ${COMPILER_ASAN_LINKER_FLAGS}")
+      set(PLATFORM_LINKFLAGS "${COMPILER_ASAN_LIBRARY} ${COMPILER_ASAN_LINKER_FLAGS}")
-      set(PLATFORM_LINKFLAGS_DEBUG "\"${COMPILER_ASAN_LIBRARY}\" \"${COMPILER_ASAN_LIBRARY_THUNK}\" ${COMPILER_ASAN_LINKER_FLAGS}")
+      set(PLATFORM_LINKFLAGS_DEBUG "${COMPILER_ASAN_LIBRARY} ${COMPILER_ASAN_LINKER_FLAGS}")
     endif()
   endif()
 endif()
@@ -1250,6 +1223,13 @@ if(WITH_OPENGL)
   add_definitions(-DWITH_OPENGL)
 endif()
 
+#-----------------------------------------------------------------------------
+# Configure Vulkan.
+
+if(WITH_VULKAN_BACKEND)
+  list(APPEND BLENDER_GL_LIBRARIES ${VULKAN_LIBRARIES})
+endif()
+
 # -----------------------------------------------------------------------------
 # Configure Metal
 

GNUmakefile

@@ -71,13 +71,6 @@ Static Source Code Checking
    * check_mypy:            Checks all Python scripts using mypy,
                             see: source/tools/check_source/check_mypy_config.py scripts which are included.
 
-Documentation Checking
-
-   * check_wiki_file_structure:
-     Check the WIKI documentation for the source-tree's file structure
-     matches Blender's source-code.
-     See: https://wiki.blender.org/wiki/Source/File_Structure
-
 Spell Checkers
    This runs the spell checker from the developer tools repositor.
 
@@ -488,10 +481,6 @@ check_smatch: .FORCE
 check_mypy: .FORCE
 	@$(PYTHON) "$(BLENDER_DIR)/source/tools/check_source/check_mypy.py"
 
-check_wiki_file_structure: .FORCE
-	@PYTHONIOENCODING=utf_8 $(PYTHON) \
-	    "$(BLENDER_DIR)/source/tools/check_wiki/check_wiki_file_structure.py"
-
 check_spelling_py: .FORCE
 	@cd "$(BUILD_DIR)" ; \
 	PYTHONIOENCODING=utf_8 $(PYTHON) \

build_files/cmake/Modules/FindMoltenVK.cmake

@@ -24,7 +24,7 @@ SET(_moltenvk_SEARCH_DIRS
 # FIXME: These finder modules typically don't use LIBDIR,
 # this should be set by `./build_files/cmake/platform/` instead.
 IF(DEFINED LIBDIR)
-  SET(_moltenvk_SEARCH_DIRS ${_moltenvk_SEARCH_DIRS} ${LIBDIR}/moltenvk)
+  SET(_moltenvk_SEARCH_DIRS ${_moltenvk_SEARCH_DIRS} ${LIBDIR}/vulkan/MoltenVK)
 ENDIF()
 
 FIND_PATH(MOLTENVK_INCLUDE_DIR

build_files/cmake/Modules/FindShaderC.cmake

@@ -1,63 +0,0 @@
-# SPDX-License-Identifier: BSD-3-Clause
-# Copyright 2023 Blender Foundation.
-
-# - Find ShaderC libraries
-# Find the ShaderC includes and libraries
-# This module defines
-#  SHADERC_INCLUDE_DIRS, where to find MoltenVK headers, Set when
-#                        SHADERC_INCLUDE_DIR is found.
-#  SHADERC_LIBRARIES, libraries to link against to use ShaderC.
-#  SHADERC_ROOT_DIR, The base directory to search for ShaderC.
-#                    This can also be an environment variable.
-#  SHADERC_FOUND, If false, do not try to use ShaderC.
-#
-
-# If SHADERC_ROOT_DIR was defined in the environment, use it.
-IF(NOT SHADERC_ROOT_DIR AND NOT $ENV{SHADERC_ROOT_DIR} STREQUAL "")
-  SET(SHADERC_ROOT_DIR $ENV{SHADERC_ROOT_DIR})
-ENDIF()
-
-SET(_shaderc_SEARCH_DIRS
-  ${SHADERC_ROOT_DIR}
-)
-
-# FIXME: These finder modules typically don't use LIBDIR,
-# this should be set by `./build_files/cmake/platform/` instead.
-IF(DEFINED LIBDIR)
-  SET(_shaderc_SEARCH_DIRS ${_shaderc_SEARCH_DIRS} ${LIBDIR}/shaderc)
-ENDIF()
-
-FIND_PATH(SHADERC_INCLUDE_DIR
-  NAMES
-    shaderc/shaderc.h
-  HINTS
-    ${_shaderc_SEARCH_DIRS}
-  PATH_SUFFIXES
-    include
-)
-
-FIND_LIBRARY(SHADERC_LIBRARY
-  NAMES
-    shaderc_combined
-  HINTS
-    ${_shaderc_SEARCH_DIRS}
-  PATH_SUFFIXES
-    lib
-)
-
-# handle the QUIETLY and REQUIRED arguments and set SHADERC_FOUND to TRUE if
-# all listed variables are TRUE
-INCLUDE(FindPackageHandleStandardArgs)
-FIND_PACKAGE_HANDLE_STANDARD_ARGS(ShaderC DEFAULT_MSG SHADERC_LIBRARY SHADERC_INCLUDE_DIR)
-
-IF(SHADERC_FOUND)
-  SET(SHADERC_LIBRARIES ${SHADERC_LIBRARY})
-  SET(SHADERC_INCLUDE_DIRS ${SHADERC_INCLUDE_DIR})
-ENDIF()
-
-MARK_AS_ADVANCED(
-  SHADERC_INCLUDE_DIR
-  SHADERC_LIBRARY
-)
-
-UNSET(_shaderc_SEARCH_DIRS)

build_files/cmake/Modules/FindVulkan.cmake

@@ -1,63 +0,0 @@
-# SPDX-License-Identifier: BSD-3-Clause
-# Copyright 2023 Blender Foundation.
-
-# - Find Vulkan libraries
-# Find the Vulkan includes and libraries
-# This module defines
-#  VULKAN_INCLUDE_DIRS, where to find Vulkan headers, Set when
-#                       VULKAN_INCLUDE_DIR is found.
-#  VULKAN_LIBRARIES, libraries to link against to use Vulkan.
-#  VULKAN_ROOT_DIR, The base directory to search for Vulkan.
-#                    This can also be an environment variable.
-#  VULKAN_FOUND, If false, do not try to use Vulkan.
-#
-
-# If VULKAN_ROOT_DIR was defined in the environment, use it.
-IF(NOT VULKAN_ROOT_DIR AND NOT $ENV{VULKAN_ROOT_DIR} STREQUAL "")
-  SET(VULKAN_ROOT_DIR $ENV{VULKAN_ROOT_DIR})
-ENDIF()
-
-SET(_vulkan_SEARCH_DIRS
-  ${VULKAN_ROOT_DIR}
-)
-
-# FIXME: These finder modules typically don't use LIBDIR,
-# this should be set by `./build_files/cmake/platform/` instead.
-IF(DEFINED LIBDIR)
-  SET(_vulkan_SEARCH_DIRS ${_vulkan_SEARCH_DIRS} ${LIBDIR}/vulkan)
-ENDIF()
-
-FIND_PATH(VULKAN_INCLUDE_DIR
-  NAMES
-    vulkan/vulkan.h
-  HINTS
-    ${_vulkan_SEARCH_DIRS}
-  PATH_SUFFIXES
-    include
-)
-
-FIND_LIBRARY(VULKAN_LIBRARY
-  NAMES
-    vulkan
-  HINTS
-    ${_vulkan_SEARCH_DIRS}
-  PATH_SUFFIXES
-    lib
-)
-
-# handle the QUIETLY and REQUIRED arguments and set VULKAN_FOUND to TRUE if
-# all listed variables are TRUE
-INCLUDE(FindPackageHandleStandardArgs)
-FIND_PACKAGE_HANDLE_STANDARD_ARGS(Vulkan DEFAULT_MSG VULKAN_LIBRARY VULKAN_INCLUDE_DIR)
-
-IF(VULKAN_FOUND)
-  SET(VULKAN_LIBRARIES ${VULKAN_LIBRARY})
-  SET(VULKAN_INCLUDE_DIRS ${VULKAN_INCLUDE_DIR})
-ENDIF()
-
-MARK_AS_ADVANCED(
-  VULKAN_INCLUDE_DIR
-  VULKAN_LIBRARY
-)
-
-UNSET(_vulkan_SEARCH_DIRS)

build_files/cmake/platform/platform_apple.cmake

@@ -97,8 +97,20 @@ add_bundled_libraries(materialx/lib)
 
 if(WITH_VULKAN_BACKEND)
   find_package(MoltenVK REQUIRED)
-  find_package(ShaderC REQUIRED)
+
-  find_package(Vulkan REQUIRED)
+  if(EXISTS ${LIBDIR}/vulkan)
+    set(VULKAN_FOUND On)
+    set(VULKAN_ROOT_DIR ${LIBDIR}/vulkan/macOS)
+    set(VULKAN_INCLUDE_DIR ${VULKAN_ROOT_DIR}/include)
+    set(VULKAN_LIBRARY ${VULKAN_ROOT_DIR}/lib/libvulkan.1.dylib)
+    set(SHADERC_LIBRARY ${VULKAN_ROOT_DIR}/lib/libshaderc_combined.a)
+
+    set(VULKAN_INCLUDE_DIRS ${VULKAN_INCLUDE_DIR} ${MOLTENVK_INCLUDE_DIRS})
+    set(VULKAN_LIBRARIES ${VULKAN_LIBRARY} ${SHADERC_LIBRARY} ${MOLTENVK_LIBRARIES})
+  else()
+    message(WARNING "Vulkan SDK was not found, disabling WITH_VULKAN_BACKEND")
+    set(WITH_VULKAN_BACKEND OFF)
+  endif()
 endif()
 
 if(WITH_OPENSUBDIV)

build_files/cmake/platform/platform_unix.cmake

@@ -111,7 +111,6 @@ find_package_wrapper(Epoxy REQUIRED)
 
 if(WITH_VULKAN_BACKEND)
   find_package_wrapper(Vulkan REQUIRED)
-  find_package_wrapper(ShaderC REQUIRED)
 endif()
 
 function(check_freetype_for_brotli)

build_files/config/pipeline_config.yaml

@@ -63,7 +63,7 @@ buildbot:
     optix:
         version: '7.3.0'
     ocloc:
-        version: '101.4032'
+        version: '101.3430'
     cmake:
         default:
             version: any

build_files/utils/make_bpy_wheel.py

@@ -24,7 +24,7 @@ import os
 import re
 import platform
 import string
-import setuptools
+import setuptools  # type: ignore
 import sys
 
 from typing import (
@@ -208,7 +208,7 @@ def main() -> None:
         return paths
 
     # Ensure this wheel is marked platform specific.
-    class BinaryDistribution(setuptools.dist.Distribution):
+    class BinaryDistribution(setuptools.dist.Distribution):  # type: ignore
         def has_ext_modules(self) -> bool:
             return True
 

build_files/utils/make_test.py

@@ -13,10 +13,10 @@ import sys
 import make_utils
 from make_utils import call
 
-# Parse arguments.
+# Parse arguments
 
 
-def parse_arguments() -> argparse.Namespace:
+def parse_arguments():
     parser = argparse.ArgumentParser()
     parser.add_argument("--ctest-command", default="ctest")
     parser.add_argument("--cmake-command", default="cmake")

extern/Eigen3/Eigen/src/Core/util/Macros.h

@@ -389,7 +389,7 @@
 
 // Does the compiler support result_of?
 #ifndef EIGEN_HAS_STD_RESULT_OF
-#if __cplusplus < 201703L && EIGEN_MAX_CPP_VER>=11 && ((__has_feature(cxx_lambdas) || (defined(__cplusplus) && __cplusplus >= 201103L && __cplusplus)))
+#if EIGEN_MAX_CPP_VER>=11 && ((__has_feature(cxx_lambdas) || (defined(__cplusplus) && __cplusplus >= 201103L)))
 #define EIGEN_HAS_STD_RESULT_OF 1
 #else
 #define EIGEN_HAS_STD_RESULT_OF 0

extern/audaspace/bindings/C/AUD_Device.cpp

@@ -221,7 +221,7 @@ AUD_API void AUD_Device_setListenerVelocity(AUD_Device* device, const float valu
 AUD_API double AUD_Device_getRate(AUD_Device* device)
 {
 	auto dev = device ? *device : DeviceManager::getDevice();
-	return dev ? dev->getSpecs().rate : 0.0;
+	return dev->getSpecs().rate;
 }
 
 AUD_API float AUD_Device_getSpeedOfSound(AUD_Device* device)

extern/audaspace/include/util/ThreadPool.h

@@ -87,17 +87,11 @@ public:
 	* \param args The arguments of the task.
 	* \return A future of the same type as the return type of the task.
 	*/
-#if __cplusplus > 201703L
-	template<class T, class... Args>
-	std::future<typename std::invoke_result<T, Args...>::type> enqueue(T&& t, Args&&... args)
-	{
-		using pkgdTask = std::packaged_task<typename std::invoke_result<T, Args...>::type()>;
-#else
 	template<class T, class... Args>
 	std::future<typename std::result_of<T(Args...)>::type> enqueue(T&& t, Args&&... args)
 	{
 		using pkgdTask = std::packaged_task<typename std::result_of<T(Args...)>::type()>;
-#endif
+
 		std::shared_ptr<pkgdTask> task = std::make_shared<pkgdTask>(std::bind(std::forward<T>(t), std::forward<Args>(args)...));
 		auto result = task->get_future();
 

extern/quadriflow/3rd/lemon-1.3.1/contrib/CMakeLists.txt

@@ -16,4 +16,3 @@ LINK_DIRECTORIES(
 
 # ADD_EXECUTABLE(myprog myprog-main.cc)
 # TARGET_LINK_LIBRARIES(myprog lemon)
-

extern/quadriflow/3rd/lemon-1.3.1/lemon/CMakeLists.txt

@@ -88,4 +88,3 @@ INSTALL(
   FILES ${CMAKE_CURRENT_BINARY_DIR}/lemon.pc
   DESTINATION lib/pkgconfig
 )
-

extern/quadriflow/CMakeLists.txt

@@ -50,18 +50,6 @@ set(INC_SYS
   ${EIGEN3_INCLUDE_DIRS}
 )
 
-#if(WITH_TBB)
-#  add_definitions(-DWITH_TBB)
-#
-#  list(APPEND INC_SYS
-#    ${TBB_INCLUDE_DIRS}
-#  )
-#
-#  list(APPEND LIB
-#    ${TBB_LIBRARIES}
-#  )
-#endif()
-
 set(SRC
   src/adjacent-matrix.cpp
   src/adjacent-matrix.hpp

extern/quadriflow/src/loader.cpp

@@ -10,7 +10,6 @@
 
 #include <fstream>
 #include <unordered_map>
-#include <functional>
 
 namespace qflow {
 
@@ -70,7 +69,7 @@ void load(const char* filename, MatrixXd& V, MatrixXi& F)
 	};
 
 	/// Hash function for obj_vertex
-	struct obj_vertexHash : std::function<size_t(obj_vertex)> {
+	struct obj_vertexHash {
 		std::size_t operator()(const obj_vertex &v) const {
 			size_t hash = std::hash<uint32_t>()(v.p);
 			hash = hash * 37 + std::hash<uint32_t>()(v.uv);

extern/quadriflow/src/optimizer.cpp

@@ -13,13 +13,6 @@
 #include "flow.hpp"
 #include "parametrizer.hpp"
 
-#ifdef max
-#undef max
-#endif
-#ifdef min
-#undef min
-#endif
-
 namespace qflow {
 
 #ifdef WITH_CUDA

extern/quadriflow/src/parametrizer-int.cpp

@@ -6,16 +6,6 @@
 #include <random>
 #include "optimizer.hpp"
 
-#include <algorithm>
-
-#ifdef max
-#undef max
-#endif
-#ifdef min
-#undef min
-#endif
-
-
 namespace qflow {
 
 

extern/quadriflow/src/parametrizer-mesh.cpp

@@ -8,16 +8,6 @@
 #include "dedge.hpp"
 #include <queue>
 
-#include <algorithm>
-
-#ifdef max
-#undef max
-#endif
-#ifdef min
-#undef min
-#endif
-
-
 namespace qflow {
 
 void Parametrizer::NormalizeMesh() {

extern/quadriflow/src/parametrizer-sing.cpp

@@ -2,14 +2,6 @@
 #include "field-math.hpp"
 #include "parametrizer.hpp"
 
-#ifdef max
-#undef max
-#endif
-#ifdef min
-#undef min
-#endif
-
-
 namespace qflow {
 
 void Parametrizer::ComputeOrientationSingularities() {

extern/quadriflow/src/parametrizer.hpp

@@ -19,13 +19,6 @@
 #include "post-solver.hpp"
 #include "serialize.hpp"
 
-#ifdef max
-#undef max
-#endif
-#ifdef min
-#undef min
-#endif
-
 namespace qflow {
 
 using namespace Eigen;

extern/quadriflow/src/subdivide.cpp

@@ -8,14 +8,6 @@
 #include "field-math.hpp"
 #include "parametrizer.hpp"
 
-#include <algorithm>
-#ifdef max
-#undef max
-#endif
-#ifdef min
-#undef min
-#endif
-
 namespace qflow {
 
 void subdivide(MatrixXi &F, MatrixXd &V, VectorXd& rho, VectorXi &V2E, VectorXi &E2E, VectorXi &boundary,

extern/vulkan_memory_allocator/CMakeLists.txt

@@ -7,7 +7,6 @@ set(INC
 
 set(INC_SYS
   ${VULKAN_INCLUDE_DIRS}
-  ${MOLTENVK_INCLUDE_DIRS}
 )
 
 set(SRC

extern/vulkan_memory_allocator/patches/remove_compilation_warning.diff

@@ -1,15 +0,0 @@
-diff --git a/extern/vulkan_memory_allocator/vk_mem_alloc.h b/extern/vulkan_memory_allocator/vk_mem_alloc.h
-index 60f572038c0..63a9994ba46 100644
---- a/extern/vulkan_memory_allocator/vk_mem_alloc.h
-+++ b/extern/vulkan_memory_allocator/vk_mem_alloc.h
-@@ -13371,8 +13371,8 @@ bool VmaDefragmentationContext_T::IncrementCounters(VkDeviceSize bytes)
-     // Early return when max found
-     if (++m_PassStats.allocationsMoved >= m_MaxPassAllocations || m_PassStats.bytesMoved >= m_MaxPassBytes)
-     {
--        VMA_ASSERT(m_PassStats.allocationsMoved == m_MaxPassAllocations ||
--            m_PassStats.bytesMoved == m_MaxPassBytes && "Exceeded maximal pass threshold!");
-+        VMA_ASSERT((m_PassStats.allocationsMoved == m_MaxPassAllocations ||
-+            m_PassStats.bytesMoved == m_MaxPassBytes) && "Exceeded maximal pass threshold!");
-         return true;
-     }
-     return false;

extern/vulkan_memory_allocator/vk_mem_alloc.h

@@ -13371,8 +13371,8 @@ bool VmaDefragmentationContext_T::IncrementCounters(VkDeviceSize bytes)
     // Early return when max found
     if (++m_PassStats.allocationsMoved >= m_MaxPassAllocations || m_PassStats.bytesMoved >= m_MaxPassBytes)
     {
-        VMA_ASSERT((m_PassStats.allocationsMoved == m_MaxPassAllocations ||
+        VMA_ASSERT(m_PassStats.allocationsMoved == m_MaxPassAllocations ||
-            m_PassStats.bytesMoved == m_MaxPassBytes) && "Exceeded maximal pass threshold!");
+            m_PassStats.bytesMoved == m_MaxPassBytes && "Exceeded maximal pass threshold!");
         return true;
     }
     return false;

intern/CMakeLists.txt

@@ -67,10 +67,6 @@ if(UNIX AND NOT APPLE)
   add_subdirectory(libc_compat)
 endif()
 
-if(WITH_INSTANT_MESHES)
-  add_subdirectory(instant-meshes)
-endif()
-
 if(UNIX AND NOT APPLE)
   # Important this comes after "ghost" as it uses includes defined by GHOST's CMake.
   if(WITH_GHOST_WAYLAND AND WITH_GHOST_WAYLAND_DYNLOAD)

intern/atomic/intern/atomic_ops_msvc.h

@@ -63,17 +63,17 @@
 /* Unsigned */
 ATOMIC_INLINE uint64_t atomic_add_and_fetch_uint64(uint64_t *p, uint64_t x)
 {
-  return (uint64_t)(InterlockedExchangeAdd64((int64_t *)p, (int64_t)x) + (int64_t)x);
+  return InterlockedExchangeAdd64((int64_t *)p, (int64_t)x) + x;
 }
 
 ATOMIC_INLINE uint64_t atomic_sub_and_fetch_uint64(uint64_t *p, uint64_t x)
 {
-  return (uint64_t)(InterlockedExchangeAdd64((int64_t *)p, -((int64_t)x)) - (int64_t)x);
+  return InterlockedExchangeAdd64((int64_t *)p, -((int64_t)x)) - x;
 }
 
 ATOMIC_INLINE uint64_t atomic_cas_uint64(uint64_t *v, uint64_t old, uint64_t _new)
 {
-  return (uint64_t)(InterlockedCompareExchange64((int64_t *)v, _new, old));
+  return InterlockedCompareExchange64((int64_t *)v, _new, old);
 }
 
 ATOMIC_INLINE uint64_t atomic_load_uint64(const uint64_t *v)
@@ -88,12 +88,12 @@ ATOMIC_INLINE void atomic_store_uint64(uint64_t *p, uint64_t v)
 
 ATOMIC_INLINE uint64_t atomic_fetch_and_add_uint64(uint64_t *p, uint64_t x)
 {
-  return (uint64_t)InterlockedExchangeAdd64((int64_t *)p, (int64_t)x);
+  return InterlockedExchangeAdd64((int64_t *)p, (int64_t)x);
 }
 
 ATOMIC_INLINE uint64_t atomic_fetch_and_sub_uint64(uint64_t *p, uint64_t x)
 {
-  return (uint64_t)InterlockedExchangeAdd64((int64_t *)p, -((int64_t)x));
+  return InterlockedExchangeAdd64((int64_t *)p, -((int64_t)x));
 }
 
 /* Signed */
@@ -137,17 +137,17 @@ ATOMIC_INLINE int64_t atomic_fetch_and_sub_int64(int64_t *p, int64_t x)
 /* Unsigned */
 ATOMIC_INLINE uint32_t atomic_add_and_fetch_uint32(uint32_t *p, uint32_t x)
 {
-  return (uint32_t)InterlockedExchangeAdd(p, x) + x;
+  return InterlockedExchangeAdd(p, x) + x;
 }
 
 ATOMIC_INLINE uint32_t atomic_sub_and_fetch_uint32(uint32_t *p, uint32_t x)
 {
-  return (uint32_t)InterlockedExchangeAdd(p, -((int32_t)x)) - x;
+  return InterlockedExchangeAdd(p, -((int32_t)x)) - x;
 }
 
 ATOMIC_INLINE uint32_t atomic_cas_uint32(uint32_t *v, uint32_t old, uint32_t _new)
 {
-  return (uint32_t)InterlockedCompareExchange((long *)v, _new, old);
+  return InterlockedCompareExchange((long *)v, _new, old);
 }
 
 ATOMIC_INLINE uint32_t atomic_load_uint32(const uint32_t *v)
@@ -162,17 +162,17 @@ ATOMIC_INLINE void atomic_store_uint32(uint32_t *p, uint32_t v)
 
 ATOMIC_INLINE uint32_t atomic_fetch_and_add_uint32(uint32_t *p, uint32_t x)
 {
-  return (uint32_t)InterlockedExchangeAdd(p, x);
+  return InterlockedExchangeAdd(p, x);
 }
 
 ATOMIC_INLINE uint32_t atomic_fetch_and_or_uint32(uint32_t *p, uint32_t x)
 {
-  return (uint32_t)InterlockedOr((long *)p, x);
+  return InterlockedOr((long *)p, x);
 }
 
 ATOMIC_INLINE uint32_t atomic_fetch_and_and_uint32(uint32_t *p, uint32_t x)
 {
-  return (uint32_t)InterlockedAnd((long *)p, x);
+  return InterlockedAnd((long *)p, x);
 }
 
 /* Signed */
@@ -259,9 +259,9 @@ ATOMIC_INLINE uint8_t atomic_fetch_and_or_uint8(uint8_t *p, uint8_t b)
 ATOMIC_INLINE int8_t atomic_fetch_and_and_int8(int8_t *p, int8_t b)
 {
 #if (LG_SIZEOF_PTR == 8 || LG_SIZEOF_INT == 8)
-  return (int8_t)InterlockedAnd8((char *)p, (char)b);
+  return InterlockedAnd8((char *)p, (char)b);
 #else
-  return (int8_t)_InterlockedAnd8((char *)p, (char)b);
+  return _InterlockedAnd8((char *)p, (char)b);
 #endif
 }
 
@@ -269,9 +269,9 @@ ATOMIC_INLINE int8_t atomic_fetch_and_and_int8(int8_t *p, int8_t b)
 ATOMIC_INLINE int8_t atomic_fetch_and_or_int8(int8_t *p, int8_t b)
 {
 #if (LG_SIZEOF_PTR == 8 || LG_SIZEOF_INT == 8)
-  return (int8_t)InterlockedOr8((char *)p, (char)b);
+  return InterlockedOr8((char *)p, (char)b);
 #else
-  return (int8_t)_InterlockedOr8((char *)p, (char)b);
+  return _InterlockedOr8((char *)p, (char)b);
 #endif
 }
 

intern/cycles/blender/addon/properties.py

@@ -1671,19 +1671,19 @@ class CyclesPreferences(bpy.types.AddonPreferences):
             elif device_type == 'HIP':
                 import sys
                 if sys.platform[:3] == "win":
-                    col.label(text="Requires AMD GPU with RDNA architecture", icon='BLANK1')
+                    col.label(text="Requires AMD GPU with Vega or RDNA architecture", icon='BLANK1')
                     col.label(text="and AMD Radeon Pro 21.Q4 driver or newer", icon='BLANK1')
                 elif sys.platform.startswith("linux"):
-                    col.label(text="Requires AMD GPU with RDNA architecture", icon='BLANK1')
+                    col.label(text="Requires AMD GPU with Vega or RDNA architecture", icon='BLANK1')
                     col.label(text="and AMD driver version 22.10 or newer", icon='BLANK1')
             elif device_type == 'ONEAPI':
                 import sys
                 if sys.platform.startswith("win"):
                     col.label(text="Requires Intel GPU with Xe-HPG architecture", icon='BLANK1')
-                    col.label(text="and Windows driver version 101.4032 or newer", icon='BLANK1')
+                    col.label(text="and Windows driver version 101.3430 or newer", icon='BLANK1')
                 elif sys.platform.startswith("linux"):
                     col.label(text="Requires Intel GPU with Xe-HPG architecture and", icon='BLANK1')
-                    col.label(text="  - intel-level-zero-gpu version 1.3.24931 or newer", icon='BLANK1')
+                    col.label(text="  - intel-level-zero-gpu version 1.3.23904 or newer", icon='BLANK1')
                     col.label(text="  - oneAPI Level-Zero Loader", icon='BLANK1')
             elif device_type == 'METAL':
                 col.label(text="Requires Apple Silicon with macOS 12.2 or newer", icon='BLANK1')

intern/cycles/blender/light.cpp

@@ -48,8 +48,6 @@ void BlenderSync::sync_light(BL::Object &b_parent,
     case BL::Light::type_SPOT: {
       BL::SpotLight b_spot_light(b_light);
       light->set_size(b_spot_light.shadow_soft_size());
-      light->set_axisu(transform_get_column(&tfm, 0));
-      light->set_axisv(transform_get_column(&tfm, 1));
       light->set_light_type(LIGHT_SPOT);
       light->set_spot_angle(b_spot_light.spot_size());
       light->set_spot_smooth(b_spot_light.spot_blend());

intern/cycles/device/hip/util.h

@@ -51,7 +51,7 @@ static inline bool hipSupportsDevice(const int hipDevId)
   hipDeviceGetAttribute(&major, hipDeviceAttributeComputeCapabilityMajor, hipDevId);
   hipDeviceGetAttribute(&minor, hipDeviceAttributeComputeCapabilityMinor, hipDevId);
 
-  return (major >= 10);
+  return (major >= 9);
 }
 
 CCL_NAMESPACE_END

intern/cycles/device/metal/device.mm

@@ -55,10 +55,6 @@ void device_metal_info(vector<DeviceInfo> &devices)
     info.denoisers = DENOISER_NONE;
     info.id = id;
 
-    if (MetalInfo::get_device_vendor(device) == METAL_GPU_AMD) {
-      info.has_light_tree = false;
-    }
-
     devices.push_back(info);
     device_index++;
   }

intern/cycles/device/oneapi/device_impl.cpp

@@ -377,7 +377,7 @@ void OneapiDevice::tex_alloc(device_texture &mem)
   generic_alloc(mem);
   generic_copy_to(mem);
 
-  /* Resize if needed. Also, in case of resize - allocate in advance for future allocations. */
+  /* Resize if needed. Also, in case of resize - allocate in advance for future allocs. */
   const uint slot = mem.slot;
   if (slot >= texture_info_.size()) {
     texture_info_.resize(slot + 128);
@@ -631,9 +631,9 @@ bool OneapiDevice::enqueue_kernel(KernelContext *kernel_context,
 /* Compute-runtime (ie. NEO) version is what gets returned by sycl/L0 on Windows
  * since Windows driver 101.3268. */
 /* The same min compute-runtime version is currently required across Windows and Linux.
- * For Windows driver 101.4032, compute-runtime version is 24931. */
+ * For Windows driver 101.3430, compute-runtime version is 23904. */
-static const int lowest_supported_driver_version_win = 1014032;
+static const int lowest_supported_driver_version_win = 1013430;
-static const int lowest_supported_driver_version_neo = 24931;
+static const int lowest_supported_driver_version_neo = 23904;
 
 int OneapiDevice::parse_driver_build_version(const sycl::device &device)
 {

intern/cycles/kernel/CMakeLists.txt

@@ -409,7 +409,6 @@ if(WITH_CYCLES_CUDA_BINARIES)
   string(REGEX REPLACE ".*release ([0-9]+)\\.([0-9]+).*" "\\2" CUDA_VERSION_MINOR "${NVCC_OUT}")
   set(CUDA_VERSION "${CUDA_VERSION_MAJOR}${CUDA_VERSION_MINOR}")
 
-
   # warn for other versions
   if((CUDA_VERSION STREQUAL "101") OR
      (CUDA_VERSION STREQUAL "102") OR
@@ -452,7 +451,6 @@ if(WITH_CYCLES_CUDA_BINARIES)
       -D CCL_NAMESPACE_BEGIN=
       -D CCL_NAMESPACE_END=
       -D NVCC
-      -allow-unsupported-compiler
       -m ${CUDA_BITS}
       -I ${CMAKE_CURRENT_SOURCE_DIR}/..
       -I ${CMAKE_CURRENT_SOURCE_DIR}/device/cuda

intern/cycles/kernel/closure/bsdf.h

@@ -170,7 +170,7 @@ ccl_device_inline int bsdf_sample(KernelGlobals kg,
     case CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID:
     case CLOSURE_BSDF_MICROFACET_GGX_REFRACTION_ID:
       label = bsdf_microfacet_ggx_sample(
-          sc, Ng, sd->wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
+          kg, sc, Ng, sd->wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
       break;
     case CLOSURE_BSDF_MICROFACET_MULTI_GGX_ID:
     case CLOSURE_BSDF_MICROFACET_MULTI_GGX_FRESNEL_ID:
@@ -185,7 +185,7 @@ ccl_device_inline int bsdf_sample(KernelGlobals kg,
     case CLOSURE_BSDF_MICROFACET_BECKMANN_ID:
     case CLOSURE_BSDF_MICROFACET_BECKMANN_REFRACTION_ID:
       label = bsdf_microfacet_beckmann_sample(
-          sc, Ng, sd->wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
+          kg, sc, Ng, sd->wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
       break;
     case CLOSURE_BSDF_ASHIKHMIN_SHIRLEY_ID:
       label = bsdf_ashikhmin_shirley_sample(
@@ -661,38 +661,4 @@ ccl_device void bsdf_blur(KernelGlobals kg, ccl_private ShaderClosure *sc, float
 #endif
 }
 
-ccl_device_inline Spectrum bsdf_albedo(ccl_private const ShaderData *sd,
-                                       ccl_private const ShaderClosure *sc)
-{
-  Spectrum albedo = sc->weight;
-  /* Some closures include additional components such as Fresnel terms that cause their albedo to
-   * be below 1. The point of this function is to return a best-effort estimation of their albedo,
-   * meaning the amount of reflected/refracted light that would be expected when illuminated by a
-   * uniform white background.
-   * This is used for the denoising albedo pass and diffuse/glossy/transmission color passes.
-   * NOTE: This should always match the sample_weight of the closure - as in, if there's an albedo
-   * adjustment in here, the sample_weight should also be reduced accordingly.
-   * TODO(lukas): Consider calling this function to determine the sample_weight? Would be a bit of
-   * extra overhead though. */
-#if defined(__SVM__) || defined(__OSL__)
-  switch (sc->type) {
-    case CLOSURE_BSDF_MICROFACET_GGX_FRESNEL_ID:
-    case CLOSURE_BSDF_MICROFACET_MULTI_GGX_FRESNEL_ID:
-    case CLOSURE_BSDF_MICROFACET_MULTI_GGX_GLASS_FRESNEL_ID:
-    case CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID:
-      albedo *= microfacet_fresnel((ccl_private const MicrofacetBsdf *)sc, sd->wi, sc->N);
-      break;
-    case CLOSURE_BSDF_PRINCIPLED_SHEEN_ID:
-      albedo *= ((ccl_private const PrincipledSheenBsdf *)sc)->avg_value;
-      break;
-    case CLOSURE_BSDF_HAIR_PRINCIPLED_ID:
-      albedo *= bsdf_principled_hair_albedo(sc);
-      break;
-    default:
-      break;
-  }
-#endif
-  return albedo;
-}
-
 CCL_NAMESPACE_END

intern/cycles/kernel/closure/bsdf_hair_principled.h

@@ -41,6 +41,11 @@ static_assert(sizeof(ShaderClosure) >= sizeof(PrincipledHairBSDF),
 static_assert(sizeof(ShaderClosure) >= sizeof(PrincipledHairExtra),
               "PrincipledHairExtra is too large!");
 
+ccl_device_inline float cos_from_sin(const float s)
+{
+  return safe_sqrtf(1.0f - s * s);
+}
+
 /* Gives the change in direction in the normal plane for the given angles and p-th-order
  * scattering. */
 ccl_device_inline float delta_phi(int p, float gamma_o, float gamma_t)

intern/cycles/kernel/closure/bsdf_microfacet.h

@@ -23,6 +23,8 @@ enum MicrofacetType {
 
 typedef struct MicrofacetExtra {
   Spectrum color, cspec0;
+  Spectrum fresnel_color;
+  float clearcoat;
 } MicrofacetExtra;
 
 typedef struct MicrofacetBsdf {
@@ -35,46 +37,39 @@ typedef struct MicrofacetBsdf {
 
 static_assert(sizeof(ShaderClosure) >= sizeof(MicrofacetBsdf), "MicrofacetBsdf is too large!");
 
-/* Beckmann VNDF importance sampling algorithm from:
+/* Beckmann and GGX microfacet importance sampling. */
- * Importance Sampling Microfacet-Based BSDFs using the Distribution of Visible Normals.
- * Eric Heitz and Eugene d'Eon, EGSR 2014.
- * https://hal.inria.fr/hal-00996995v2/document */
 
-ccl_device_forceinline float3 microfacet_beckmann_sample_vndf(const float3 wi,
+ccl_device_inline void microfacet_beckmann_sample_slopes(KernelGlobals kg,
-                                                              const float alpha_x,
+                                                         const float cos_theta_i,
-                                                              const float alpha_y,
+                                                         const float sin_theta_i,
-                                                              const float randu,
+                                                         float randu,
-                                                              const float randv)
+                                                         float randv,
+                                                         ccl_private float *slope_x,
+                                                         ccl_private float *slope_y,
+                                                         ccl_private float *G1i)
 {
-  /* 1. stretch wi */
-  float3 wi_ = make_float3(alpha_x * wi.x, alpha_y * wi.y, wi.z);
-  wi_ = normalize(wi_);
-
-  /* 2. sample P22_{wi}(x_slope, y_slope, 1, 1) */
-  float slope_x, slope_y;
-  float cos_phi_i = 1.0f;
-  float sin_phi_i = 0.0f;
-
-  if (wi_.z >= 0.99999f) {
   /* Special case (normal incidence). */
+  if (cos_theta_i >= 0.99999f) {
     const float r = sqrtf(-logf(randu));
     const float phi = M_2PI_F * randv;
-    slope_x = r * cosf(phi);
+    *slope_x = r * cosf(phi);
-    slope_y = r * sinf(phi);
+    *slope_y = r * sinf(phi);
+    *G1i = 1.0f;
+
+    return;
   }
-  else {
+
   /* Precomputations. */
-    const float cos_theta_i = wi_.z;
-    const float sin_theta_i = sin_from_cos(cos_theta_i);
   const float tan_theta_i = sin_theta_i / cos_theta_i;
+  const float inv_a = tan_theta_i;
   const float cot_theta_i = 1.0f / tan_theta_i;
   const float erf_a = fast_erff(cot_theta_i);
   const float exp_a2 = expf(-cot_theta_i * cot_theta_i);
   const float SQRT_PI_INV = 0.56418958354f;
+  const float Lambda = 0.5f * (erf_a - 1.0f) + (0.5f * SQRT_PI_INV) * (exp_a2 * inv_a);
+  const float G1 = 1.0f / (1.0f + Lambda); /* masking */
 
-    float invlen = 1.0f / sin_theta_i;
+  *G1i = G1;
-    cos_phi_i = wi_.x * invlen;
-    sin_phi_i = wi_.y * invlen;
 
   /* Based on paper from Wenzel Jakob
    * An Improved Visible Normal Sampling Routine for the Beckmann Distribution
@@ -121,13 +116,111 @@ ccl_device_forceinline float3 microfacet_beckmann_sample_vndf(const float3 wi,
     current.y = 1.0f + current.x + K * expf(-sqr(inv_erf)) - y_exact;
   }
 
-    slope_x = inv_erf;
+  *slope_x = inv_erf;
-    slope_y = fast_ierff(2.0f * randv - 1.0f);
+  *slope_y = fast_ierff(2.0f * randv - 1.0f);
+}
+
+/* GGX microfacet importance sampling from:
+ *
+ * Importance Sampling Microfacet-Based BSDFs using the Distribution of Visible Normals.
+ * E. Heitz and E. d'Eon, EGSR 2014
+ */
+
+ccl_device_inline void microfacet_ggx_sample_slopes(const float cos_theta_i,
+                                                    const float sin_theta_i,
+                                                    float randu,
+                                                    float randv,
+                                                    ccl_private float *slope_x,
+                                                    ccl_private float *slope_y,
+                                                    ccl_private float *G1i)
+{
+  /* Special case (normal incidence). */
+  if (cos_theta_i >= 0.99999f) {
+    const float r = sqrtf(randu / (1.0f - randu));
+    const float phi = M_2PI_F * randv;
+    *slope_x = r * cosf(phi);
+    *slope_y = r * sinf(phi);
+    *G1i = 1.0f;
+
+    return;
+  }
+
+  /* Precomputations. */
+  const float tan_theta_i = sin_theta_i / cos_theta_i;
+  const float G1_inv = 0.5f * (1.0f + safe_sqrtf(1.0f + tan_theta_i * tan_theta_i));
+
+  *G1i = 1.0f / G1_inv;
+
+  /* Sample slope_x. */
+  const float A = 2.0f * randu * G1_inv - 1.0f;
+  const float AA = A * A;
+  const float tmp = 1.0f / (AA - 1.0f);
+  const float B = tan_theta_i;
+  const float BB = B * B;
+  const float D = safe_sqrtf(BB * (tmp * tmp) - (AA - BB) * tmp);
+  const float slope_x_1 = B * tmp - D;
+  const float slope_x_2 = B * tmp + D;
+  *slope_x = (A < 0.0f || slope_x_2 * tan_theta_i > 1.0f) ? slope_x_1 : slope_x_2;
+
+  /* Sample slope_y. */
+  float S;
+
+  if (randv > 0.5f) {
+    S = 1.0f;
+    randv = 2.0f * (randv - 0.5f);
+  }
+  else {
+    S = -1.0f;
+    randv = 2.0f * (0.5f - randv);
+  }
+
+  const float z = (randv * (randv * (randv * 0.27385f - 0.73369f) + 0.46341f)) /
+                  (randv * (randv * (randv * 0.093073f + 0.309420f) - 1.000000f) + 0.597999f);
+  *slope_y = S * z * safe_sqrtf(1.0f + (*slope_x) * (*slope_x));
+}
+
+template<MicrofacetType m_type>
+ccl_device_forceinline float3 microfacet_sample_stretched(KernelGlobals kg,
+                                                          const float3 wi,
+                                                          const float alpha_x,
+                                                          const float alpha_y,
+                                                          const float randu,
+                                                          const float randv,
+                                                          ccl_private float *G1i)
+{
+  /* 1. stretch wi */
+  float3 wi_ = make_float3(alpha_x * wi.x, alpha_y * wi.y, wi.z);
+  wi_ = normalize(wi_);
+
+  /* Compute polar coordinates of wi_. */
+  float costheta_ = 1.0f;
+  float sintheta_ = 0.0f;
+  float cosphi_ = 1.0f;
+  float sinphi_ = 0.0f;
+
+  if (wi_.z < 0.99999f) {
+    costheta_ = wi_.z;
+    sintheta_ = safe_sqrtf(1.0f - costheta_ * costheta_);
+
+    float invlen = 1.0f / sintheta_;
+    cosphi_ = wi_.x * invlen;
+    sinphi_ = wi_.y * invlen;
+  }
+
+  /* 2. sample P22_{wi}(x_slope, y_slope, 1, 1) */
+  float slope_x, slope_y;
+
+  if (m_type == MicrofacetType::BECKMANN) {
+    microfacet_beckmann_sample_slopes(
+        kg, costheta_, sintheta_, randu, randv, &slope_x, &slope_y, G1i);
+  }
+  else {
+    microfacet_ggx_sample_slopes(costheta_, sintheta_, randu, randv, &slope_x, &slope_y, G1i);
   }
 
   /* 3. rotate */
-  float tmp = cos_phi_i * slope_x - sin_phi_i * slope_y;
+  float tmp = cosphi_ * slope_x - sinphi_ * slope_y;
-  slope_y = sin_phi_i * slope_x + cos_phi_i * slope_y;
+  slope_y = sinphi_ * slope_x + cosphi_ * slope_y;
   slope_x = tmp;
 
   /* 4. unstretch */
@@ -138,43 +231,6 @@ ccl_device_forceinline float3 microfacet_beckmann_sample_vndf(const float3 wi,
   return normalize(make_float3(-slope_x, -slope_y, 1.0f));
 }
 
-/* GGX VNDF importance sampling algorithm from:
- * Sampling the GGX Distribution of Visible Normals.
- * Eric Heitz, JCGT Vol. 7, No. 4, 2018.
- * https://jcgt.org/published/0007/04/01/ */
-ccl_device_forceinline float3 microfacet_ggx_sample_vndf(const float3 wi,
-                                                         const float alpha_x,
-                                                         const float alpha_y,
-                                                         const float randu,
-                                                         const float randv)
-{
-  /* Section 3.2: Transforming the view direction to the hemisphere configuration. */
-  float3 wi_ = normalize(make_float3(alpha_x * wi.x, alpha_y * wi.y, wi.z));
-
-  /* Section 4.1: Orthonormal basis. */
-  float lensq = sqr(wi_.x) + sqr(wi_.y);
-  float3 T1, T2;
-  if (lensq > 1e-7f) {
-    T1 = make_float3(-wi_.y, wi_.x, 0.0f) * inversesqrtf(lensq);
-    T2 = cross(wi_, T1);
-  }
-  else {
-    /* Normal incidence, any basis is fine. */
-    T1 = make_float3(1.0f, 0.0f, 0.0f);
-    T2 = make_float3(0.0f, 1.0f, 0.0f);
-  }
-
-  /* Section 4.2: Parameterization of the projected area. */
-  float2 t = concentric_sample_disk(randu, randv);
-  t.y = mix(safe_sqrtf(1.0f - sqr(t.x)), t.y, 0.5f * (1.0f + wi_.z));
-
-  /* Section 4.3: Reprojection onto hemisphere. */
-  float3 H_ = t.x * T1 + t.y * T2 + safe_sqrtf(1.0f - len_squared(t)) * wi_;
-
-  /* Section 3.4: Transforming the normal back to the ellipsoid configuration. */
-  return normalize(make_float3(alpha_x * H_.x, alpha_y * H_.y, max(0.0f, H_.z)));
-}
-
 /* Calculate the reflection color
  *
  * If fresnel is used, the color is an interpolation of the F0 color and white
@@ -182,25 +238,26 @@ ccl_device_forceinline float3 microfacet_ggx_sample_vndf(const float3 wi,
  *
  * Else it is simply white
  */
-ccl_device_forceinline Spectrum microfacet_fresnel(ccl_private const MicrofacetBsdf *bsdf,
+ccl_device_forceinline Spectrum reflection_color(ccl_private const MicrofacetBsdf *bsdf,
-                                                   float3 wi,
+                                                 float3 L,
                                                  float3 H)
 {
-  if (CLOSURE_IS_BSDF_MICROFACET_FRESNEL(bsdf->type)) {
+  Spectrum F = one_spectrum();
-    return interpolate_fresnel_color(wi, H, bsdf->ior, bsdf->extra->cspec0);
-  }
-  else if (bsdf->type == CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID) {
-    return make_spectrum(fresnel_dielectric_cos(dot(wi, H), bsdf->ior));
-  }
-  else {
-    return one_spectrum();
-  }
-}
 
-ccl_device_forceinline void bsdf_microfacet_adjust_weight(ccl_private const ShaderData *sd,
+  bool use_clearcoat = bsdf->type == CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID;
-                                                          ccl_private MicrofacetBsdf *bsdf)
+  bool use_fresnel = (bsdf->type == CLOSURE_BSDF_MICROFACET_GGX_FRESNEL_ID || use_clearcoat);
-{
+
-  bsdf->sample_weight *= average(microfacet_fresnel(bsdf, sd->wi, bsdf->N));
+  if (use_fresnel) {
+    float F0 = fresnel_dielectric_cos(1.0f, bsdf->ior);
+
+    F = interpolate_fresnel_color(L, H, bsdf->ior, F0, bsdf->extra->cspec0);
+  }
+
+  if (use_clearcoat) {
+    F *= 0.25f * bsdf->extra->clearcoat;
+  }
+
+  return F;
 }
 
 /* Generalized Trowbridge-Reitz for clearcoat. */
@@ -214,48 +271,37 @@ ccl_device_forceinline float bsdf_clearcoat_D(float alpha2, float cos_NH)
   return (alpha2 - 1.0f) / (M_PI_F * logf(alpha2) * t);
 }
 
-/* Smith shadowing-masking term, here in the non-separable form.
+/* Monodirectional shadowing-masking term. */
- * For details, see:
- * Understanding the Masking-Shadowing Function in Microfacet-Based BRDFs.
- * Eric Heitz, JCGT Vol. 3, No. 2, 2014.
- * https://jcgt.org/published/0003/02/03/ */
 template<MicrofacetType m_type>
-ccl_device_inline float bsdf_lambda_from_sqr_alpha_tan_n(float sqr_alpha_tan_n)
+ccl_device_inline float bsdf_G1_from_sqr_alpha_tan_n(float sqr_alpha_tan_n)
 {
   if (m_type == MicrofacetType::GGX) {
-    /* Equation 72. */
+    return 2.0f / (1.0f + sqrtf(1.0f + sqr_alpha_tan_n));
-    return 0.5f * (sqrtf(1.0f + sqr_alpha_tan_n) - 1.0f);
   }
   else {
-    /* m_type == MicrofacetType::BECKMANN
+    /* m_type == MicrofacetType::BECKMANN */
-     * Approximation from below Equation 69. */
-    if (sqr_alpha_tan_n < 0.39f) {
-      /* Equivalent to a >= 1.6f, but also handles sqr_alpha_tan_n == 0.0f cleanly. */
-      return 0.0f;
-    }
-
     const float a = inversesqrtf(sqr_alpha_tan_n);
-    return ((0.396f * a - 1.259f) * a + 1.0f) / ((2.181f * a + 3.535f) * a);
+    return (a > 1.6f) ? 1.0f : ((2.181f * a + 3.535f) * a) / ((2.577f * a + 2.276f) * a + 1.0f);
   }
 }
 
-template<MicrofacetType m_type> ccl_device_inline float bsdf_lambda(float alpha2, float cos_N)
+template<MicrofacetType m_type> ccl_device_inline float bsdf_G1(float alpha2, float cos_N)
 {
-  return bsdf_lambda_from_sqr_alpha_tan_n<m_type>(alpha2 * fmaxf(1.0f / sqr(cos_N) - 1.0f, 0.0f));
+  return bsdf_G1_from_sqr_alpha_tan_n<m_type>(alpha2 * fmaxf(1.0f / (cos_N * cos_N) - 1.0f, 0.0f));
 }
 
 template<MicrofacetType m_type>
-ccl_device_inline float bsdf_aniso_lambda(float alpha_x, float alpha_y, float3 V)
+ccl_device_inline float bsdf_aniso_G1(float alpha_x, float alpha_y, float3 V)
 {
-  const float sqr_alpha_tan_n = (sqr(alpha_x * V.x) + sqr(alpha_y * V.y)) / sqr(V.z);
+  return bsdf_G1_from_sqr_alpha_tan_n<m_type>((sqr(alpha_x * V.x) + sqr(alpha_y * V.y)) /
-  return bsdf_lambda_from_sqr_alpha_tan_n<m_type>(sqr_alpha_tan_n);
+                                              sqr(V.z));
 }
 
-/* Combined shadowing-masking term. */
+/* Smith's separable shadowing-masking term. */
 template<MicrofacetType m_type>
 ccl_device_inline float bsdf_G(float alpha2, float cos_NI, float cos_NO)
 {
-  return 1.0f / (1.0f + bsdf_lambda<m_type>(alpha2, cos_NI) + bsdf_lambda<m_type>(alpha2, cos_NO));
+  return bsdf_G1<m_type>(alpha2, cos_NI) * bsdf_G1<m_type>(alpha2, cos_NO);
 }
 
 /* Normal distribution function. */
@@ -289,6 +335,22 @@ ccl_device_inline float bsdf_aniso_D(float alpha_x, float alpha_y, float3 H)
   }
 }
 
+ccl_device_forceinline void bsdf_microfacet_fresnel_color(ccl_private const ShaderData *sd,
+                                                          ccl_private MicrofacetBsdf *bsdf)
+{
+  kernel_assert(CLOSURE_IS_BSDF_MICROFACET_FRESNEL(bsdf->type));
+
+  float F0 = fresnel_dielectric_cos(1.0f, bsdf->ior);
+  bsdf->extra->fresnel_color = interpolate_fresnel_color(
+      sd->wi, bsdf->N, bsdf->ior, F0, bsdf->extra->cspec0);
+
+  if (bsdf->type == CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID) {
+    bsdf->extra->fresnel_color *= 0.25f * bsdf->extra->clearcoat;
+  }
+
+  bsdf->sample_weight *= average(bsdf->extra->fresnel_color);
+}
+
 template<MicrofacetType m_type>
 ccl_device Spectrum bsdf_microfacet_eval(ccl_private const ShaderClosure *sc,
                                          const float3 Ng,
@@ -320,7 +382,7 @@ ccl_device Spectrum bsdf_microfacet_eval(ccl_private const ShaderClosure *sc,
   H *= inv_len_H;
 
   const float cos_NH = dot(N, H);
-  float D, lambdaI, lambdaO;
+  float D, G1i, G1o;
 
   /* TODO: add support for anisotropic transmission. */
   if (alpha_x == alpha_y || m_refractive) { /* Isotropic. */
@@ -337,8 +399,8 @@ ccl_device Spectrum bsdf_microfacet_eval(ccl_private const ShaderClosure *sc,
       D = bsdf_D<m_type>(alpha2, cos_NH);
     }
 
-    lambdaI = bsdf_lambda<m_type>(alpha2, cos_NI);
+    G1i = bsdf_G1<m_type>(alpha2, cos_NI);
-    lambdaO = bsdf_lambda<m_type>(alpha2, cos_NO);
+    G1o = bsdf_G1<m_type>(alpha2, cos_NO);
   }
   else { /* Anisotropic. */
     float3 X, Y;
@@ -350,23 +412,25 @@ ccl_device Spectrum bsdf_microfacet_eval(ccl_private const ShaderClosure *sc,
 
     D = bsdf_aniso_D<m_type>(alpha_x, alpha_y, local_H);
 
-    lambdaI = bsdf_aniso_lambda<m_type>(alpha_x, alpha_y, local_I);
+    G1i = bsdf_aniso_G1<m_type>(alpha_x, alpha_y, local_I);
-    lambdaO = bsdf_aniso_lambda<m_type>(alpha_x, alpha_y, local_O);
+    G1o = bsdf_aniso_G1<m_type>(alpha_x, alpha_y, local_O);
   }
 
-  const float common = D / cos_NI *
+  const float common = G1i * D / cos_NI *
                        (m_refractive ?
                             sqr(bsdf->ior * inv_len_H) * fabsf(dot(H, wi) * dot(H, wo)) :
                             0.25f);
 
-  *pdf = common / (1.0f + lambdaI);
+  *pdf = common;
 
-  const Spectrum F = microfacet_fresnel(bsdf, wo, H);
+  const Spectrum F = m_refractive ? one_spectrum() : reflection_color(bsdf, wo, H);
-  return F * common / (1.0f + lambdaO + lambdaI);
+
+  return F * G1o * common;
 }
 
 template<MicrofacetType m_type>
-ccl_device int bsdf_microfacet_sample(ccl_private const ShaderClosure *sc,
+ccl_device int bsdf_microfacet_sample(KernelGlobals kg,
+                                      ccl_private const ShaderClosure *sc,
                                       float3 Ng,
                                       float3 wi,
                                       float randu,
@@ -402,15 +466,10 @@ ccl_device int bsdf_microfacet_sample(ccl_private const ShaderClosure *sc,
 
   /* Importance sampling with distribution of visible normals. Vectors are transformed to local
    * space before and after sampling. */
+  float G1i;
   const float3 local_I = make_float3(dot(X, wi), dot(Y, wi), cos_NI);
-  float3 local_H;
+  const float3 local_H = microfacet_sample_stretched<m_type>(
-  if (m_type == MicrofacetType::GGX) {
+      kg, local_I, alpha_x, alpha_y, randu, randv, &G1i);
-    local_H = microfacet_ggx_sample_vndf(local_I, alpha_x, alpha_y, randu, randv);
-  }
-  else {
-    /* m_type == MicrofacetType::BECKMANN */
-    local_H = microfacet_beckmann_sample_vndf(local_I, alpha_x, alpha_y, randu, randv);
-  }
 
   const float3 H = X * local_H.x + Y * local_H.y + N * local_H.z;
   const float cos_NH = local_H.z;
@@ -443,12 +502,19 @@ ccl_device int bsdf_microfacet_sample(ccl_private const ShaderClosure *sc,
     label |= LABEL_SINGULAR;
     /* Some high number for MIS. */
     *pdf = 1e6f;
-    *eval = make_spectrum(1e6f) * microfacet_fresnel(bsdf, *wo, H);
+    *eval = make_spectrum(1e6f);
+
+    bool use_fresnel = (bsdf->type == CLOSURE_BSDF_MICROFACET_GGX_FRESNEL_ID ||
+                        bsdf->type == CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID);
+
+    if (use_fresnel && !m_refractive) {
+      *eval *= reflection_color(bsdf, *wo, H);
+    }
   }
   else {
     label |= LABEL_GLOSSY;
     float cos_NO = dot(N, *wo);
-    float D, lambdaI, lambdaO;
+    float D, G1o;
 
     /* TODO: add support for anisotropic transmission. */
     if (alpha_x == alpha_y || m_refractive) { /* Isotropic. */
@@ -460,32 +526,34 @@ ccl_device int bsdf_microfacet_sample(ccl_private const ShaderClosure *sc,
         /* The masking-shadowing term for clearcoat has a fixed alpha of 0.25
          * => alpha2 = 0.25 * 0.25 */
         alpha2 = 0.0625f;
+
+        /* Recalculate G1i. */
+        G1i = bsdf_G1<m_type>(alpha2, cos_NI);
       }
       else {
         D = bsdf_D<m_type>(alpha2, cos_NH);
       }
 
-      lambdaO = bsdf_lambda<m_type>(alpha2, cos_NO);
+      G1o = bsdf_G1<m_type>(alpha2, cos_NO);
-      lambdaI = bsdf_lambda<m_type>(alpha2, cos_NI);
     }
     else { /* Anisotropic. */
       const float3 local_O = make_float3(dot(X, *wo), dot(Y, *wo), cos_NO);
 
       D = bsdf_aniso_D<m_type>(alpha_x, alpha_y, local_H);
 
-      lambdaO = bsdf_aniso_lambda<m_type>(alpha_x, alpha_y, local_O);
+      G1o = bsdf_aniso_G1<m_type>(alpha_x, alpha_y, local_O);
-      lambdaI = bsdf_aniso_lambda<m_type>(alpha_x, alpha_y, local_I);
     }
 
     const float cos_HO = dot(H, *wo);
-    const float common = D / cos_NI *
+    const float common = G1i * D / cos_NI *
                          (m_refractive ? fabsf(cos_HI * cos_HO) / sqr(cos_HO + cos_HI / m_eta) :
                                          0.25f);
 
-    *pdf = common / (1.0f + lambdaI);
+    *pdf = common;
 
-    Spectrum F = microfacet_fresnel(bsdf, *wo, H);
+    Spectrum F = m_refractive ? one_spectrum() : reflection_color(bsdf, *wo, H);
-    *eval = F * common / (1.0f + lambdaI + lambdaO);
+
+    *eval = G1o * common * F;
   }
 
   *sampled_roughness = make_float2(alpha_x, alpha_y);
@@ -519,6 +587,14 @@ ccl_device int bsdf_microfacet_ggx_setup(ccl_private MicrofacetBsdf *bsdf)
   return SD_BSDF | SD_BSDF_HAS_EVAL;
 }
 
+/* Required to maintain OSL interface. */
+ccl_device int bsdf_microfacet_ggx_isotropic_setup(ccl_private MicrofacetBsdf *bsdf)
+{
+  bsdf->alpha_y = bsdf->alpha_x;
+
+  return bsdf_microfacet_ggx_setup(bsdf);
+}
+
 ccl_device int bsdf_microfacet_ggx_fresnel_setup(ccl_private MicrofacetBsdf *bsdf,
                                                  ccl_private const ShaderData *sd)
 {
@@ -529,7 +605,7 @@ ccl_device int bsdf_microfacet_ggx_fresnel_setup(ccl_private MicrofacetBsdf *bsd
 
   bsdf->type = CLOSURE_BSDF_MICROFACET_GGX_FRESNEL_ID;
 
-  bsdf_microfacet_adjust_weight(sd, bsdf);
+  bsdf_microfacet_fresnel_color(sd, bsdf);
 
   return SD_BSDF | SD_BSDF_HAS_EVAL;
 }
@@ -537,12 +613,14 @@ ccl_device int bsdf_microfacet_ggx_fresnel_setup(ccl_private MicrofacetBsdf *bsd
 ccl_device int bsdf_microfacet_ggx_clearcoat_setup(ccl_private MicrofacetBsdf *bsdf,
                                                    ccl_private const ShaderData *sd)
 {
+  bsdf->extra->cspec0 = saturate(bsdf->extra->cspec0);
+
   bsdf->alpha_x = saturatef(bsdf->alpha_x);
   bsdf->alpha_y = bsdf->alpha_x;
 
   bsdf->type = CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID;
 
-  bsdf_microfacet_adjust_weight(sd, bsdf);
+  bsdf_microfacet_fresnel_color(sd, bsdf);
 
   return SD_BSDF | SD_BSDF_HAS_EVAL;
 }
@@ -576,7 +654,8 @@ ccl_device Spectrum bsdf_microfacet_ggx_eval(ccl_private const ShaderClosure *sc
   return bsdf_microfacet_eval<MicrofacetType::GGX>(sc, Ng, wi, wo, pdf);
 }
 
-ccl_device int bsdf_microfacet_ggx_sample(ccl_private const ShaderClosure *sc,
+ccl_device int bsdf_microfacet_ggx_sample(KernelGlobals kg,
+                                          ccl_private const ShaderClosure *sc,
                                           float3 Ng,
                                           float3 wi,
                                           float randu,
@@ -588,7 +667,7 @@ ccl_device int bsdf_microfacet_ggx_sample(ccl_private const ShaderClosure *sc,
                                           ccl_private float *eta)
 {
   return bsdf_microfacet_sample<MicrofacetType::GGX>(
-      sc, Ng, wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
+      kg, sc, Ng, wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
 }
 
 /* Beckmann microfacet with Smith shadow-masking from:
@@ -605,6 +684,14 @@ ccl_device int bsdf_microfacet_beckmann_setup(ccl_private MicrofacetBsdf *bsdf)
   return SD_BSDF | SD_BSDF_HAS_EVAL;
 }
 
+/* Required to maintain OSL interface. */
+ccl_device int bsdf_microfacet_beckmann_isotropic_setup(ccl_private MicrofacetBsdf *bsdf)
+{
+  bsdf->alpha_y = bsdf->alpha_x;
+
+  return bsdf_microfacet_beckmann_setup(bsdf);
+}
+
 ccl_device int bsdf_microfacet_beckmann_refraction_setup(ccl_private MicrofacetBsdf *bsdf)
 {
   bsdf->alpha_x = saturatef(bsdf->alpha_x);
@@ -631,7 +718,8 @@ ccl_device Spectrum bsdf_microfacet_beckmann_eval(ccl_private const ShaderClosur
   return bsdf_microfacet_eval<MicrofacetType::BECKMANN>(sc, Ng, wi, wo, pdf);
 }
 
-ccl_device int bsdf_microfacet_beckmann_sample(ccl_private const ShaderClosure *sc,
+ccl_device int bsdf_microfacet_beckmann_sample(KernelGlobals kg,
+                                               ccl_private const ShaderClosure *sc,
                                                float3 Ng,
                                                float3 wi,
                                                float randu,
@@ -643,7 +731,7 @@ ccl_device int bsdf_microfacet_beckmann_sample(ccl_private const ShaderClosure *
                                                ccl_private float *eta)
 {
   return bsdf_microfacet_sample<MicrofacetType::BECKMANN>(
-      sc, Ng, wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
+      kg, sc, Ng, wi, randu, randv, eval, wo, pdf, sampled_roughness, eta);
 }
 
 CCL_NAMESPACE_END

intern/cycles/kernel/closure/bsdf_microfacet_multi.h

@@ -43,7 +43,7 @@ ccl_device_forceinline float2 mf_sampleP22_11(const float cosI,
     return make_float2(r * cosf(phi), r * sinf(phi));
   }
 
-  const float sinI = sin_from_cos(cosI);
+  const float sinI = safe_sqrtf(1.0f - cosI * cosI);
   const float tanI = sinI / cosI;
   const float projA = 0.5f * (cosI + 1.0f);
   if (projA < 0.0001f)
@@ -401,7 +401,7 @@ ccl_device int bsdf_microfacet_multi_ggx_fresnel_setup(ccl_private MicrofacetBsd
 
   bsdf->type = CLOSURE_BSDF_MICROFACET_MULTI_GGX_FRESNEL_ID;
 
-  bsdf_microfacet_adjust_weight(sd, bsdf);
+  bsdf_microfacet_fresnel_color(sd, bsdf);
 
   return bsdf_microfacet_multi_ggx_common_setup(bsdf);
 }
@@ -575,7 +575,7 @@ ccl_device int bsdf_microfacet_multi_ggx_glass_fresnel_setup(ccl_private Microfa
 
   bsdf->type = CLOSURE_BSDF_MICROFACET_MULTI_GGX_GLASS_FRESNEL_ID;
 
-  bsdf_microfacet_adjust_weight(sd, bsdf);
+  bsdf_microfacet_fresnel_color(sd, bsdf);
 
   return SD_BSDF | SD_BSDF_HAS_EVAL | SD_BSDF_NEEDS_LCG;
 }

intern/cycles/kernel/closure/bsdf_microfacet_multi_impl.h

@@ -73,8 +73,9 @@ ccl_device_forceinline Spectrum MF_FUNCTION_FULL_NAME(mf_eval)(float3 wi,
   eval = make_spectrum(val);
 #endif
 
+  float F0 = fresnel_dielectric_cos(1.0f, eta);
   if (use_fresnel) {
-    throughput = interpolate_fresnel_color(wi, wh, eta, cspec0);
+    throughput = interpolate_fresnel_color(wi, wh, eta, F0, cspec0);
 
     eval *= throughput;
   }
@@ -143,11 +144,11 @@ ccl_device_forceinline Spectrum MF_FUNCTION_FULL_NAME(mf_eval)(float3 wi,
         throughput *= color;
       }
       else if (use_fresnel && order > 0) {
-        throughput *= interpolate_fresnel_color(wi_prev, wm, eta, cspec0);
+        throughput *= interpolate_fresnel_color(wi_prev, wm, eta, F0, cspec0);
       }
 #else /* MF_MULTI_GLOSSY */
       if (use_fresnel && order > 0) {
-        throughput *= interpolate_fresnel_color(-wr, wm, eta, cspec0);
+        throughput *= interpolate_fresnel_color(-wr, wm, eta, F0, cspec0);
       }
       wr = mf_sample_phase_glossy(-wr, &throughput, wm);
 #endif
@@ -191,6 +192,8 @@ ccl_device_forceinline Spectrum MF_FUNCTION_FULL_NAME(mf_sample)(float3 wi,
   float G1_r = 0.0f;
   bool outside = true;
 
+  float F0 = fresnel_dielectric_cos(1.0f, eta);
+
   int order;
   for (order = 0; order < 10; order++) {
     /* Sample microfacet height. */
@@ -226,12 +229,22 @@ ccl_device_forceinline Spectrum MF_FUNCTION_FULL_NAME(mf_sample)(float3 wi,
         throughput *= color;
       }
       else {
-        throughput *= interpolate_fresnel_color(wi_prev, wm, eta, cspec0);
+        Spectrum t_color = interpolate_fresnel_color(wi_prev, wm, eta, F0, cspec0);
+
+        if (order == 0)
+          throughput = t_color;
+        else
+          throughput *= t_color;
       }
     }
 #else /* MF_MULTI_GLOSSY */
     if (use_fresnel) {
-      throughput *= interpolate_fresnel_color(-wr, wm, eta, cspec0);
+      Spectrum t_color = interpolate_fresnel_color(-wr, wm, eta, F0, cspec0);
+
+      if (order == 0)
+        throughput = t_color;
+      else
+        throughput *= t_color;
     }
     wr = mf_sample_phase_glossy(-wr, &throughput, wm);
 #endif

intern/cycles/kernel/closure/bsdf_util.h

@@ -89,21 +89,19 @@ ccl_device float schlick_fresnel(float u)
   return m2 * m2 * m;  // pow(m, 5)
 }
 
-/* Calculate the fresnel color, which is a blend between white and the F0 color */
+/* Calculate the fresnel color which is a blend between white and the F0 color (cspec0) */
-ccl_device_forceinline Spectrum interpolate_fresnel_color(float3 L,
+ccl_device_forceinline Spectrum
-                                                          float3 H,
+interpolate_fresnel_color(float3 L, float3 H, float ior, float F0, Spectrum cspec0)
-                                                          float ior,
-                                                          Spectrum F0)
 {
-  /* Compute the real Fresnel term and remap it from real_F0..1 to F0..1.
+  /* Calculate the fresnel interpolation factor
-   * The reason why we use this remapping instead of directly doing the
+   * The value from fresnel_dielectric_cos(...) has to be normalized because
-   * Schlick approximation lerp(F0, 1.0, (1.0-cosLH)^5) is that for cases
+   * the cspec0 keeps the F0 color
-   * with similar IORs (e.g. ice in water), the relative IOR can be close
+   */
-   * enough to 1.0 that the Schlick approximation becomes inaccurate. */
+  float F0_norm = 1.0f / (1.0f - F0);
-  float real_F = fresnel_dielectric_cos(dot(L, H), ior);
+  float FH = (fresnel_dielectric_cos(dot(L, H), ior) - F0) * F0_norm;
-  float real_F0 = fresnel_dielectric_cos(1.0f, ior);
 
-  return mix(F0, one_spectrum(), inverse_lerp(real_F0, 1.0f, real_F));
+  /* Blend between white and a specular color with respect to the fresnel */
+  return cspec0 * (1.0f - FH) + make_spectrum(FH);
 }
 
 ccl_device float3 ensure_valid_reflection(float3 Ng, float3 I, float3 N)

intern/cycles/kernel/closure/volume.h

@@ -88,7 +88,7 @@ henyey_greenstrein_sample(float3 D, float g, float randu, float randv, ccl_priva
     }
   }
 
-  float sin_theta = sin_from_cos(cos_theta);
+  float sin_theta = safe_sqrtf(1.0f - cos_theta * cos_theta);
   float phi = M_2PI_F * randv;
   float3 dir = make_float3(sin_theta * cosf(phi), sin_theta * sinf(phi), cos_theta);
 

intern/cycles/kernel/device/oneapi/compat.h

@@ -195,15 +195,7 @@ using sycl::half;
 #define fmodf(x, y) sycl::fmod((x), (y))
 #define lgammaf(x) sycl::lgamma((x))
 
-/* `sycl::native::cos` precision is not sufficient and `-ffast-math` lets
+#define cosf(x) sycl::native::cos(((float)(x)))
- * the current DPC++ compiler overload `sycl::cos` with it.
- * We work around this issue by directly calling the SPIRV implementation which
- * provides greater precision. */
-#if defined(__SYCL_DEVICE_ONLY__) && defined(__SPIR__)
-#  define cosf(x) __spirv_ocl_cos(((float)(x)))
-#else
-#  define cosf(x) sycl::cos(((float)(x)))
-#endif
 #define sinf(x) sycl::native::sin(((float)(x)))
 #define powf(x, y) sycl::native::powr(((float)(x)), ((float)(y)))
 #define tanf(x) sycl::native::tan(((float)(x)))

intern/cycles/kernel/film/denoising_passes.h

@@ -58,7 +58,23 @@ ccl_device_forceinline void film_write_denoising_features_surface(KernelGlobals
     normal += sc->N * sc->sample_weight;
     sum_weight += sc->sample_weight;
 
-    if (sc->type == CLOSURE_BSDF_PRINCIPLED_DIFFUSE_ID) {
+    Spectrum closure_albedo = sc->weight;
+    /* Closures that include a Fresnel term typically have weights close to 1 even though their
+     * actual contribution is significantly lower.
+     * To account for this, we scale their weight by the average fresnel factor (the same is also
+     * done for the sample weight in the BSDF setup, so we don't need to scale that here). */
+    if (CLOSURE_IS_BSDF_MICROFACET_FRESNEL(sc->type)) {
+      ccl_private MicrofacetBsdf *bsdf = (ccl_private MicrofacetBsdf *)sc;
+      closure_albedo *= bsdf->extra->fresnel_color;
+    }
+    else if (sc->type == CLOSURE_BSDF_PRINCIPLED_SHEEN_ID) {
+      ccl_private PrincipledSheenBsdf *bsdf = (ccl_private PrincipledSheenBsdf *)sc;
+      closure_albedo *= bsdf->avg_value;
+    }
+    else if (sc->type == CLOSURE_BSDF_HAIR_PRINCIPLED_ID) {
+      closure_albedo *= bsdf_principled_hair_albedo(sc);
+    }
+    else if (sc->type == CLOSURE_BSDF_PRINCIPLED_DIFFUSE_ID) {
       /* BSSRDF already accounts for weight, retro-reflection would double up. */
       ccl_private const PrincipledDiffuseBsdf *bsdf = (ccl_private const PrincipledDiffuseBsdf *)
           sc;
@@ -67,7 +83,6 @@ ccl_device_forceinline void film_write_denoising_features_surface(KernelGlobals
       }
     }
 
-    Spectrum closure_albedo = bsdf_albedo(sd, sc);
     if (bsdf_get_specular_roughness_squared(sc) > sqr(0.075f)) {
       diffuse_albedo += closure_albedo;
       sum_nonspecular_weight += sc->sample_weight;

intern/cycles/kernel/geom/curve_intersect.h

@@ -720,7 +720,7 @@ ccl_device_inline void curve_shader_setup(KernelGlobals kg,
     const float3 tangent = normalize(dPdu);
     const float3 bitangent = normalize(cross(tangent, -D));
     const float sine = sd->v;
-    const float cosine = cos_from_sin(sine);
+    const float cosine = safe_sqrtf(1.0f - sine * sine);
 
     sd->N = normalize(sine * bitangent - cosine * normalize(cross(tangent, bitangent)));
 #  if 0

intern/cycles/kernel/integrator/mnee.h

@@ -704,9 +704,9 @@ ccl_device_forceinline bool mnee_compute_transfer_matrix(ccl_private const Shade
     float ilo = -eta * ilh;
 
     float cos_theta = dot(wo, m.n);
-    float sin_theta = sin_from_cos(cos_theta);
+    float sin_theta = safe_sqrtf(1.f - sqr(cos_theta));
     float cos_phi = dot(wo, s);
-    float sin_phi = sin_from_cos(cos_phi);
+    float sin_phi = safe_sqrtf(1.f - sqr(cos_phi));
 
     /* Wo = (cos_phi * sin_theta) * s + (sin_phi * sin_theta) * t + cos_theta * n. */
     float3 dH_dtheta = ilo * (cos_theta * (cos_phi * s + sin_phi * t) - sin_theta * m.n);

intern/cycles/kernel/integrator/subsurface_random_walk.h

@@ -136,7 +136,7 @@ ccl_device_forceinline float diffusion_length_dwivedi(float alpha)
 
 ccl_device_forceinline float3 direction_from_cosine(float3 D, float cos_theta, float randv)
 {
-  float sin_theta = sin_from_cos(cos_theta);
+  float sin_theta = safe_sqrtf(1.0f - cos_theta * cos_theta);
   float phi = M_2PI_F * randv;
   float3 dir = make_float3(sin_theta * cosf(phi), sin_theta * sinf(phi), cos_theta);
 

intern/cycles/kernel/integrator/surface_shader.h

@@ -621,7 +621,7 @@ ccl_device Spectrum surface_shader_diffuse(KernelGlobals kg, ccl_private const S
     ccl_private const ShaderClosure *sc = &sd->closure[i];
 
     if (CLOSURE_IS_BSDF_DIFFUSE(sc->type) || CLOSURE_IS_BSSRDF(sc->type))
-      eval += bsdf_albedo(sd, sc);
+      eval += sc->weight;
   }
 
   return eval;
@@ -635,7 +635,7 @@ ccl_device Spectrum surface_shader_glossy(KernelGlobals kg, ccl_private const Sh
     ccl_private const ShaderClosure *sc = &sd->closure[i];
 
     if (CLOSURE_IS_BSDF_GLOSSY(sc->type))
-      eval += bsdf_albedo(sd, sc);
+      eval += sc->weight;
   }
 
   return eval;
@@ -649,7 +649,7 @@ ccl_device Spectrum surface_shader_transmission(KernelGlobals kg, ccl_private co
     ccl_private const ShaderClosure *sc = &sd->closure[i];
 
     if (CLOSURE_IS_BSDF_TRANSMISSION(sc->type))
-      eval += bsdf_albedo(sd, sc);
+      eval += sc->weight;
   }
 
   return eval;

intern/cycles/kernel/light/area.h

@@ -102,7 +102,7 @@ ccl_device float area_light_spread_attenuation(const float3 D,
     /* The factor M_PI_F comes from integrating the radiance over the hemisphere */
     return (cos_a > 0.9999997f) ? M_PI_F : 0.0f;
   }
-  const float sin_a = sin_from_cos(cos_a);
+  const float sin_a = safe_sqrtf(1.0f - sqr(cos_a));
   const float tan_a = sin_a / cos_a;
   return max((tan_half_spread - tan_a) * normalize_spread, 0.0f);
 }

intern/cycles/kernel/light/spot.h

@@ -7,13 +7,24 @@
 
 CCL_NAMESPACE_BEGIN
 
-ccl_device float spot_light_attenuation(const ccl_global KernelSpotLight *spot, float3 ray)
+ccl_device float spot_light_attenuation(float3 dir,
+                                        float cos_half_spot_angle,
+                                        float spot_smooth,
+                                        float3 N)
 {
-  const float3 scaled_ray = safe_normalize(
+  float attenuation = dot(dir, N);
-      make_float3(dot(ray, spot->axis_u), dot(ray, spot->axis_v), dot(ray, spot->dir)) /
-      spot->len);
 
-  return smoothstepf((scaled_ray.z - spot->cos_half_spot_angle) / spot->spot_smooth);
+  if (attenuation <= cos_half_spot_angle) {
+    attenuation = 0.0f;
+  }
+  else {
+    float t = attenuation - cos_half_spot_angle;
+
+    if (t < spot_smooth && spot_smooth != 0.0f)
+      attenuation *= smoothstepf(t / spot_smooth);
+  }
+
+  return attenuation;
 }
 
 template<bool in_volume_segment>
@@ -46,7 +57,8 @@ ccl_device_inline bool spot_light_sample(const ccl_global KernelLight *klight,
   ls->eval_fac = (0.25f * M_1_PI_F) * invarea;
 
   /* spot light attenuation */
-  ls->eval_fac *= spot_light_attenuation(&klight->spot, -ls->D);
+  ls->eval_fac *= spot_light_attenuation(
+      klight->spot.dir, klight->spot.cos_half_spot_angle, klight->spot.spot_smooth, -ls->D);
   if (!in_volume_segment && ls->eval_fac == 0.0f) {
     return false;
   }
@@ -75,7 +87,8 @@ ccl_device_forceinline void spot_light_update_position(const ccl_global KernelLi
   ls->pdf = invarea;
 
   /* spot light attenuation */
-  ls->eval_fac *= spot_light_attenuation(&klight->spot, ls->Ng);
+  ls->eval_fac *= spot_light_attenuation(
+      klight->spot.dir, klight->spot.cos_half_spot_angle, klight->spot.spot_smooth, ls->Ng);
 }
 
 ccl_device_inline bool spot_light_intersect(const ccl_global KernelLight *klight,
@@ -116,7 +129,8 @@ ccl_device_inline bool spot_light_sample_from_intersection(
   ls->pdf = invarea;
 
   /* spot light attenuation */
-  ls->eval_fac *= spot_light_attenuation(&klight->spot, -ls->D);
+  ls->eval_fac *= spot_light_attenuation(
+      klight->spot.dir, klight->spot.cos_half_spot_angle, klight->spot.spot_smooth, -ls->D);
 
   if (ls->eval_fac == 0.0f) {
     return false;

intern/cycles/kernel/light/tree.h

@@ -47,6 +47,11 @@ ccl_device float light_tree_cos_bounding_box_angle(const BoundingBox bbox,
   return cos_theta_u;
 }
 
+ccl_device_forceinline float sin_from_cos(const float c)
+{
+  return safe_sqrtf(1.0f - sqr(c));
+}
+
 /* Compute vector v as in Fig .8. P_v is the corresponding point along the ray. */
 ccl_device float3 compute_v(
     const float3 centroid, const float3 P, const float3 D, const float3 bcone_axis, const float t)

intern/cycles/kernel/light/triangle.h

@@ -218,7 +218,7 @@ ccl_device_forceinline bool triangle_light_sample(KernelGlobals kg,
     /* Finally, select a random point along the edge of the new triangle
      * That point on the spherical triangle is the sampled ray direction */
     const float z = 1.0f - randv * (1.0f - dot(C_, B));
-    ls->D = z * B + sin_from_cos(z) * safe_normalize(C_ - dot(C_, B) * B);
+    ls->D = z * B + safe_sqrtf(1.0f - z * z) * safe_normalize(C_ - dot(C_, B) * B);
 
     /* calculate intersection with the planar triangle */
     if (!ray_triangle_intersect(

intern/cycles/kernel/osl/closures_setup.h

@@ -209,8 +209,15 @@ ccl_device void osl_closure_microfacet_setup(KernelGlobals kg,
   if (closure->distribution == make_string("ggx", 11253504724482777663ull) ||
       closure->distribution == make_string("default", 4430693559278735917ull)) {
     if (!closure->refract) {
+      if (closure->alpha_x == closure->alpha_y) {
+        /* Isotropic */
+        sd->flag |= bsdf_microfacet_ggx_isotropic_setup(bsdf);
+      }
+      else {
+        /* Anisotropic */
         sd->flag |= bsdf_microfacet_ggx_setup(bsdf);
       }
+    }
     else {
       sd->flag |= bsdf_microfacet_ggx_refraction_setup(bsdf);
     }
@@ -218,8 +225,15 @@ ccl_device void osl_closure_microfacet_setup(KernelGlobals kg,
   /* Beckmann */
   else {
     if (!closure->refract) {
+      if (closure->alpha_x == closure->alpha_y) {
+        /* Isotropic */
+        sd->flag |= bsdf_microfacet_beckmann_isotropic_setup(bsdf);
+      }
+      else {
+        /* Anisotropic */
         sd->flag |= bsdf_microfacet_beckmann_setup(bsdf);
       }
+    }
     else {
       sd->flag |= bsdf_microfacet_beckmann_refraction_setup(bsdf);
     }
@@ -244,9 +258,9 @@ ccl_device void osl_closure_microfacet_ggx_setup(
   }
 
   bsdf->N = ensure_valid_reflection(sd->Ng, sd->wi, closure->N);
-  bsdf->alpha_x = bsdf->alpha_y = closure->alpha_x;
+  bsdf->alpha_x = closure->alpha_x;
 
-  sd->flag |= bsdf_microfacet_ggx_setup(bsdf);
+  sd->flag |= bsdf_microfacet_ggx_isotropic_setup(bsdf);
 }
 
 ccl_device void osl_closure_microfacet_ggx_aniso_setup(
@@ -331,6 +345,7 @@ ccl_device void osl_closure_microfacet_ggx_fresnel_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = rgb_to_spectrum(closure->cspec0);
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = zero_float3();
 
@@ -368,6 +383,7 @@ ccl_device void osl_closure_microfacet_ggx_aniso_fresnel_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = rgb_to_spectrum(closure->cspec0);
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = closure->T;
 
@@ -410,6 +426,7 @@ ccl_device void osl_closure_microfacet_multi_ggx_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = zero_spectrum();
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = zero_float3();
 
@@ -450,6 +467,7 @@ ccl_device void osl_closure_microfacet_multi_ggx_glass_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = zero_spectrum();
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = zero_float3();
 
@@ -490,6 +508,7 @@ ccl_device void osl_closure_microfacet_multi_ggx_aniso_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = zero_spectrum();
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = closure->T;
 
@@ -532,6 +551,7 @@ ccl_device void osl_closure_microfacet_multi_ggx_fresnel_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = rgb_to_spectrum(closure->cspec0);
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = zero_float3();
 
@@ -572,6 +592,7 @@ ccl_device void osl_closure_microfacet_multi_ggx_glass_fresnel_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = rgb_to_spectrum(closure->cspec0);
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = zero_float3();
 
@@ -612,6 +633,7 @@ ccl_device void osl_closure_microfacet_multi_ggx_aniso_fresnel_setup(
   bsdf->extra = extra;
   bsdf->extra->color = rgb_to_spectrum(closure->color);
   bsdf->extra->cspec0 = rgb_to_spectrum(closure->cspec0);
+  bsdf->extra->clearcoat = 0.0f;
 
   bsdf->T = closure->T;
 
@@ -638,9 +660,9 @@ ccl_device void osl_closure_microfacet_beckmann_setup(
   }
 
   bsdf->N = ensure_valid_reflection(sd->Ng, sd->wi, closure->N);
-  bsdf->alpha_x = bsdf->alpha_y = closure->alpha_x;
+  bsdf->alpha_x = closure->alpha_x;
 
-  sd->flag |= bsdf_microfacet_beckmann_setup(bsdf);
+  sd->flag |= bsdf_microfacet_beckmann_isotropic_setup(bsdf);
 }
 
 ccl_device void osl_closure_microfacet_beckmann_aniso_setup(
@@ -843,18 +865,27 @@ ccl_device void osl_closure_principled_clearcoat_setup(
     float3 weight,
     ccl_private const PrincipledClearcoatClosure *closure)
 {
-  weight *= 0.25f * closure->clearcoat;
   ccl_private MicrofacetBsdf *bsdf = (ccl_private MicrofacetBsdf *)bsdf_alloc(
       sd, sizeof(MicrofacetBsdf), rgb_to_spectrum(weight));
   if (!bsdf) {
     return;
   }
 
+  MicrofacetExtra *extra = (MicrofacetExtra *)closure_alloc_extra(sd, sizeof(MicrofacetExtra));
+  if (!extra) {
+    return;
+  }
+
   bsdf->N = ensure_valid_reflection(sd->Ng, sd->wi, closure->N);
   bsdf->alpha_x = closure->clearcoat_roughness;
   bsdf->alpha_y = closure->clearcoat_roughness;
   bsdf->ior = 1.5f;
 
+  bsdf->extra = extra;
+  bsdf->extra->color = zero_spectrum();
+  bsdf->extra->cspec0 = make_spectrum(0.04f);
+  bsdf->extra->clearcoat = closure->clearcoat;
+
   bsdf->T = zero_float3();
 
   sd->flag |= bsdf_microfacet_ggx_clearcoat_setup(bsdf, sd);

intern/cycles/kernel/osl/osl.h

@@ -161,10 +161,7 @@ ccl_device_inline void osl_eval_nodes(KernelGlobals kg,
                         /* shadeindex = */ 0);
 #  endif
 
-  if constexpr (type == SHADER_TYPE_DISPLACEMENT) {
+  if (globals.Ci) {
-    sd->P = globals.P;
-  }
-  else if (globals.Ci) {
     flatten_closure_tree(kg, sd, path_flag, globals.Ci);
   }
 }

intern/cycles/kernel/sample/mapping.h

@@ -67,18 +67,17 @@ ccl_device_inline void sample_uniform_cone(const float3 N,
                                            ccl_private float3 *wo,
                                            ccl_private float *pdf)
 {
-  const float cosThetaMin = cosf(angle);
+  float zMin = cosf(angle);
-  const float cosTheta = mix(cosThetaMin, 1.0f, randu);
+  float z = zMin - zMin * randu + randu;
-  const float sinTheta = sin_from_cos(cosTheta);
+  float r = safe_sqrtf(1.0f - sqr(z));
-  const float phi = M_2PI_F * randv;
+  float phi = M_2PI_F * randv;
-  const float x = sinTheta * cosf(phi);
+  float x = r * cosf(phi);
-  const float y = sinTheta * sinf(phi);
+  float y = r * sinf(phi);
-  const float z = cosTheta;
 
   float3 T, B;
   make_orthonormals(N, &T, &B);
   *wo = x * T + y * B + z * N;
-  *pdf = M_1_2PI_F / (1.0f - cosThetaMin);
+  *pdf = M_1_2PI_F / (1.0f - zMin);
 }
 
 ccl_device_inline float pdf_uniform_cone(const float3 N, float3 D, float angle)

intern/cycles/kernel/svm/brick.h

@@ -46,8 +46,17 @@ ccl_device_noinline_cpu float2 svm_brick(float3 p,
   float tint = saturatef((brick_noise((rownum << 16) + (bricknum & 0xFFFF)) + bias));
   float min_dist = min(min(x, y), min(brick_width - x, row_height - y));
 
+  float mortar;
+  if (min_dist >= mortar_size) {
+    mortar = 0.0f;
+  }
+  else if (mortar_smooth == 0.0f) {
+    mortar = 1.0f;
+  }
+  else {
     min_dist = 1.0f - min_dist / mortar_size;
-  float mortar = smoothstepf(min_dist / mortar_smooth);
+    mortar = (min_dist < mortar_smooth) ? smoothstepf(min_dist / mortar_smooth) : 1.0f;
+  }
 
   return make_float2(tint, mortar);
 }

intern/cycles/kernel/svm/closure.h

@@ -333,6 +333,7 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
             bsdf->extra->cspec0 = rgb_to_spectrum(
                 (specular * 0.08f * tmp_col) * (1.0f - metallic) + base_color * metallic);
             bsdf->extra->color = rgb_to_spectrum(base_color);
+            bsdf->extra->clearcoat = 0.0f;
 
             /* setup bsdf */
             if (distribution == CLOSURE_BSDF_MICROFACET_GGX_GLASS_ID ||
@@ -382,6 +383,7 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
 
                 bsdf->extra->color = rgb_to_spectrum(base_color);
                 bsdf->extra->cspec0 = rgb_to_spectrum(cspec0);
+                bsdf->extra->clearcoat = 0.0f;
 
                 /* setup bsdf */
                 sd->flag |= bsdf_microfacet_ggx_fresnel_setup(bsdf, sd);
@@ -438,6 +440,7 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
 
               bsdf->extra->color = rgb_to_spectrum(base_color);
               bsdf->extra->cspec0 = rgb_to_spectrum(cspec0);
+              bsdf->extra->clearcoat = 0.0f;
 
               /* setup bsdf */
               sd->flag |= bsdf_microfacet_multi_ggx_glass_fresnel_setup(bsdf, sd);
@@ -452,21 +455,31 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
 #ifdef __CAUSTICS_TRICKS__
       if (kernel_data.integrator.caustics_reflective || (path_flag & PATH_RAY_DIFFUSE) == 0) {
 #endif
-        Spectrum clearcoat_weight = 0.25f * clearcoat * weight;
+        if (clearcoat > CLOSURE_WEIGHT_CUTOFF) {
           ccl_private MicrofacetBsdf *bsdf = (ccl_private MicrofacetBsdf *)bsdf_alloc(
-            sd, sizeof(MicrofacetBsdf), clearcoat_weight);
+              sd, sizeof(MicrofacetBsdf), weight);
+          ccl_private MicrofacetExtra *extra =
+              (bsdf != NULL) ?
+                  (ccl_private MicrofacetExtra *)closure_alloc_extra(sd, sizeof(MicrofacetExtra)) :
+                  NULL;
 
-        if (bsdf) {
+          if (bsdf && extra) {
             bsdf->N = clearcoat_normal;
             bsdf->T = zero_float3();
             bsdf->ior = 1.5f;
+            bsdf->extra = extra;
 
             bsdf->alpha_x = clearcoat_roughness * clearcoat_roughness;
             bsdf->alpha_y = clearcoat_roughness * clearcoat_roughness;
 
+            bsdf->extra->color = zero_spectrum();
+            bsdf->extra->cspec0 = make_spectrum(0.04f);
+            bsdf->extra->clearcoat = clearcoat;
+
             /* setup bsdf */
             sd->flag |= bsdf_microfacet_ggx_clearcoat_setup(bsdf, sd);
           }
+        }
 #ifdef __CAUSTICS_TRICKS__
       }
 #endif
@@ -571,6 +584,7 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
         if (bsdf->extra) {
           bsdf->extra->color = rgb_to_spectrum(stack_load_float3(stack, data_node.w));
           bsdf->extra->cspec0 = zero_spectrum();
+          bsdf->extra->clearcoat = 0.0f;
           sd->flag |= bsdf_microfacet_multi_ggx_setup(bsdf);
         }
       }
@@ -710,6 +724,7 @@ ccl_device_noinline int svm_node_closure_bsdf(KernelGlobals kg,
       kernel_assert(stack_valid(data_node.z));
       bsdf->extra->color = rgb_to_spectrum(stack_load_float3(stack, data_node.z));
       bsdf->extra->cspec0 = zero_spectrum();
+      bsdf->extra->clearcoat = 0.0f;
 
       /* setup bsdf */
       sd->flag |= bsdf_microfacet_multi_ggx_glass_setup(bsdf);

intern/cycles/kernel/svm/types.h

@@ -489,7 +489,8 @@ typedef enum ClosureType {
 #define CLOSURE_IS_BSDF_MICROFACET_FRESNEL(type) \
   (type == CLOSURE_BSDF_MICROFACET_MULTI_GGX_FRESNEL_ID || \
    type == CLOSURE_BSDF_MICROFACET_MULTI_GGX_GLASS_FRESNEL_ID || \
-   type == CLOSURE_BSDF_MICROFACET_GGX_FRESNEL_ID)
+   type == CLOSURE_BSDF_MICROFACET_GGX_FRESNEL_ID || \
+   type == CLOSURE_BSDF_MICROFACET_GGX_CLEARCOAT_ID)
 #define CLOSURE_IS_BSDF_OR_BSSRDF(type) (type <= CLOSURE_BSSRDF_RANDOM_WALK_FIXED_RADIUS_ID)
 #define CLOSURE_IS_BSSRDF(type) \
   (type >= CLOSURE_BSSRDF_BURLEY_ID && type <= CLOSURE_BSSRDF_RANDOM_WALK_FIXED_RADIUS_ID)

intern/cycles/kernel/types.h

@@ -74,8 +74,7 @@ CCL_NAMESPACE_BEGIN
 #define __VOLUME__
 
 /* TODO: solve internal compiler errors and enable light tree on HIP. */
-/* TODO: solve internal compiler perf issue and enable light tree on Metal/AMD. */
+#ifdef __KERNEL_HIP__
-#if defined(__KERNEL_HIP__) || defined(__KERNEL_METAL_AMD__)
 #  undef __LIGHT_TREE__
 #endif
 
@@ -1291,14 +1290,12 @@ typedef struct KernelCurveSegment {
 static_assert_align(KernelCurveSegment, 8);
 
 typedef struct KernelSpotLight {
-  packed_float3 axis_u;
   float radius;
-  packed_float3 axis_v;
   float invarea;
-  packed_float3 dir;
   float cos_half_spot_angle;
-  packed_float3 len;
   float spot_smooth;
+  packed_float3 dir;
+  float pad;
 } KernelSpotLight;
 
 /* PointLight is SpotLight with only radius and invarea being used. */

intern/cycles/scene/geometry.cpp

@@ -23,10 +23,7 @@
 #include "subd/patch_table.h"
 #include "subd/split.h"
 
-#ifdef WITH_OSL
+#include "kernel/osl/globals.h"
-#  include "kernel/osl/globals.h"
-#  include "kernel/osl/services.h"
-#endif
 
 #include "util/foreach.h"
 #include "util/log.h"
@@ -309,11 +306,6 @@ void GeometryManager::update_osl_globals(Device *device, Scene *scene)
 {
 #ifdef WITH_OSL
   OSLGlobals *og = (OSLGlobals *)device->get_cpu_osl_memory();
-  if (og == nullptr) {
-    /* Can happen when rendering with multiple GPUs, but no CPU (in which case the name maps filled
-     * below are not used anyway) */
-    return;
-  }
 
   og->object_name_map.clear();
   og->object_names.clear();
@@ -1674,7 +1666,6 @@ void GeometryManager::device_update_displacement_images(Device *device,
   TaskPool pool;
   ImageManager *image_manager = scene->image_manager;
   set<int> bump_images;
-  bool has_osl_node = false;
   foreach (Geometry *geom, scene->geometry) {
     if (geom->is_modified()) {
       /* Geometry-level check for hair shadow transparency.
@@ -1694,9 +1685,6 @@ void GeometryManager::device_update_displacement_images(Device *device,
           continue;
         }
         foreach (ShaderNode *node, shader->graph->nodes) {
-          if (node->special_type == SHADER_SPECIAL_TYPE_OSL) {
-            has_osl_node = true;
-          }
           if (node->special_type != SHADER_SPECIAL_TYPE_IMAGE_SLOT) {
             continue;
           }
@@ -1712,28 +1700,6 @@ void GeometryManager::device_update_displacement_images(Device *device,
       }
     }
   }
-
-#ifdef WITH_OSL
-  /* If any OSL node is used for displacement, it may reference a texture. But it's
-   * unknown which ones, so have to load them all. */
-  if (has_osl_node) {
-    set<OSLRenderServices *> services_shared;
-    device->foreach_device([&services_shared](Device *sub_device) {
-      OSLGlobals *og = (OSLGlobals *)sub_device->get_cpu_osl_memory();
-      services_shared.insert(og->services);
-    });
-
-    for (OSLRenderServices *services : services_shared) {
-      for (auto it = services->textures.begin(); it != services->textures.end(); ++it) {
-        if (it->second->handle.get_manager() == image_manager) {
-          const int slot = it->second->handle.svm_slot();
-          bump_images.insert(slot);
-        }
-      }
-    }
-  }
-#endif
-
   foreach (int slot, bump_images) {
     pool.push(function_bind(
         &ImageManager::device_update_slot, image_manager, device, scene, slot, &progress));

intern/cycles/scene/light.cpp

@@ -1076,31 +1076,23 @@ void LightManager::device_update_lights(Device *device, DeviceScene *dscene, Sce
     else if (light->light_type == LIGHT_SPOT) {
       shader_id &= ~SHADER_AREA_LIGHT;
 
-      float3 len;
-      float3 axis_u = normalize_len(light->axisu, &len.x);
-      float3 axis_v = normalize_len(light->axisv, &len.y);
-      float3 dir = normalize_len(light->dir, &len.z);
-      if (len.z == 0.0f) {
-        dir = zero_float3();
-      }
-
       float radius = light->size;
       float invarea = (radius > 0.0f) ? 1.0f / (M_PI_F * radius * radius) : 1.0f;
       float cos_half_spot_angle = cosf(light->spot_angle * 0.5f);
       float spot_smooth = (1.0f - cos_half_spot_angle) * light->spot_smooth;
+      float3 dir = light->dir;
+
+      dir = safe_normalize(dir);
 
       if (light->use_mis && radius > 0.0f)
         shader_id |= SHADER_USE_MIS;
 
       klights[light_index].co = co;
-      klights[light_index].spot.axis_u = axis_u;
       klights[light_index].spot.radius = radius;
-      klights[light_index].spot.axis_v = axis_v;
       klights[light_index].spot.invarea = invarea;
-      klights[light_index].spot.dir = dir;
       klights[light_index].spot.cos_half_spot_angle = cos_half_spot_angle;
-      klights[light_index].spot.len = len;
       klights[light_index].spot.spot_smooth = spot_smooth;
+      klights[light_index].spot.dir = dir;
     }
 
     klights[light_index].shader_id = shader_id;

intern/cycles/scene/light_tree.cpp

@@ -156,13 +156,7 @@ LightTreePrimitive::LightTreePrimitive(Scene *scene, int prim_id, int object_id)
     }
     else if (type == LIGHT_SPOT) {
       bcone.theta_o = 0;
-
+      bcone.theta_e = lamp->get_spot_angle() * 0.5f;
-      const float unscaled_theta_e = lamp->get_spot_angle() * 0.5f;
-      const float len_u = len(lamp->get_axisu());
-      const float len_v = len(lamp->get_axisv());
-      const float len_w = len(lamp->get_dir());
-
-      bcone.theta_e = fast_atanf(fast_tanf(unscaled_theta_e) * fmaxf(len_u, len_v) / len_w);
 
       /* Point and spot lights can emit light from any point within its radius. */
       const float3 radius = make_float3(size);

intern/cycles/scene/osl.cpp

@@ -1241,7 +1241,6 @@ void OSLCompiler::compile(OSLGlobals *og, Shader *shader)
 
     shader->has_surface = false;
     shader->has_surface_transparent = false;
-    shader->has_surface_raytrace = false;
     shader->has_surface_bssrdf = false;
     shader->has_bump = has_bump;
     shader->has_bssrdf_bump = has_bump;

intern/cycles/util/math.h

@@ -483,12 +483,6 @@ ccl_device_inline float compatible_signf(float f)
 
 ccl_device_inline float smoothstepf(float f)
 {
-  if (f <= 0.0f) {
-    return 0.0f;
-  }
-  if (f >= 1.0f) {
-    return 1.0f;
-  }
   float ff = f * f;
   return (3.0f * ff - 2.0f * ff * f);
 }
@@ -756,16 +750,6 @@ ccl_device_inline float sqr(float a)
   return a * a;
 }
 
-ccl_device_inline float sin_from_cos(const float c)
-{
-  return safe_sqrtf(1.0f - sqr(c));
-}
-
-ccl_device_inline float cos_from_sin(const float s)
-{
-  return safe_sqrtf(1.0f - sqr(s));
-}
-
 ccl_device_inline float pow20(float a)
 {
   return sqr(sqr(sqr(sqr(a)) * a));

intern/cycles/util/math_float2.h

@@ -134,11 +134,6 @@ ccl_device_inline float len(const float2 a)
   return sqrtf(dot(a, a));
 }
 
-ccl_device_inline float len_squared(const float2 a)
-{
-  return dot(a, a);
-}
-
 #if !defined(__KERNEL_METAL__)
 ccl_device_inline float distance(const float2 a, const float2 b)
 {

intern/ghost/CMakeLists.txt

@@ -85,12 +85,10 @@ if(WITH_VULKAN_BACKEND)
 
   list(APPEND INC_SYS
     ${VULKAN_INCLUDE_DIRS}
-    ${MOLTENVK_INCLUDE_DIRS}
   )
 
   list(APPEND LIB
     ${VULKAN_LIBRARIES}
-    ${MOLTENVK_LIBRARIES}
   )
 
   add_definitions(-DWITH_VULKAN_BACKEND)
@@ -596,7 +594,4 @@ if(WITH_XR_OPENXR)
   unset(XR_PLATFORM_DEFINES)
 endif()
 
-remove_cc_flag("-fsanitize=address")
-remove_cc_flag("/fsanitize=address")
-
 blender_add_lib(bf_intern_ghost "${SRC}" "${INC}" "${INC_SYS}" "${LIB}")

intern/ghost/GHOST_C-api.h

@@ -1201,7 +1201,7 @@ void GHOST_GetVulkanHandles(GHOST_ContextHandle context,
                             void *r_instance,
                             void *r_physical_device,
                             void *r_device,
-                            uint32_t *r_graphic_queue_family);
+                            uint32_t *r_graphic_queue_familly);
 
 /**
  * Return VULKAN back-buffer resources handles for the given window.

intern/ghost/GHOST_Types.h

@@ -9,18 +9,6 @@
 
 #include <stdint.h>
 
-#ifndef ATTR_NO_OPT
-#  ifdef __clang__
-#    define ATTR_NO_OPT __attribute__((optnone))
-#  elif defined(_MSC_VER)
-#    define ATTR_NO_OPT __pragma(optimize("", off))
-#  elif defined(__GNUC__)
-#    define ATTR_NO_OPT __attribute__((optimize("O0")))
-#  else
-#    define ATTR_NO_OPT
-#  endif
-#endif
-
 #ifdef WITH_CXX_GUARDEDALLOC
 #  include "MEM_guardedalloc.h"
 #else

intern/ghost/intern/GHOST_C-api.cpp

@@ -1203,10 +1203,10 @@ void GHOST_GetVulkanHandles(GHOST_ContextHandle contexthandle,
                             void *r_instance,
                             void *r_physical_device,
                             void *r_device,
-                            uint32_t *r_graphic_queue_family)
+                            uint32_t *r_graphic_queue_familly)
 {
   GHOST_IContext *context = (GHOST_IContext *)contexthandle;
-  context->getVulkanHandles(r_instance, r_physical_device, r_device, r_graphic_queue_family);
+  context->getVulkanHandles(r_instance, r_physical_device, r_device, r_graphic_queue_familly);
 }
 
 void GHOST_GetVulkanBackbuffer(GHOST_WindowHandle windowhandle,

intern/ghost/intern/GHOST_Context.h

@@ -15,14 +15,6 @@
 
 #include <cstdlib>  // for NULL
 
-#if defined(__clang__) || defined(__GCC__)
-#  define ATTR_NO_ASAN __attribute__((no_sanitize("address")))
-#elif _MSC_VER
-#  define ATTR_NO_ASAN __declspec(no_sanitize_address)
-#else
-#  define ATTR_NO_ASAN
-#endif
-
 class GHOST_Context : public GHOST_IContext {
  public:
   /**
@@ -150,7 +142,7 @@ class GHOST_Context : public GHOST_IContext {
   virtual GHOST_TSuccess getVulkanHandles(void * /*r_instance*/,
                                           void * /*r_physical_device*/,
                                           void * /*r_device*/,
-                                          uint32_t * /*r_graphic_queue_family*/) override
+                                          uint32_t * /*r_graphic_queue_familly*/) override
   {
     return GHOST_kFailure;
   };

intern/ghost/intern/GHOST_ContextCGL.h

@@ -23,22 +23,6 @@
 @class NSView;
 
 class GHOST_ContextCGL : public GHOST_Context {
-
- public:
-  /* Defines the number of simultaneous command buffers which can be in flight.
-   * The default limit of `64` is considered to be optimal for Blender. Too many command buffers
-   * will result in workload fragmentation and additional system-level overhead. This limit should
-   * also only be increased if the application is consistently exceeding the limit, and there are
-   * no command buffer leaks.
-   *
-   * If this limit is reached, starting a new command buffer will fail. The Metal back-end will
-   * therefore stall until completion and log a warning when this limit is reached in order to
-   * ensure correct function of the app.
-   *
-   * It is generally preferable to reduce the prevalence of GPU_flush or GPU Context switches
-   * (which will both break command submissions), rather than increasing this limit. */
-  static const int max_command_buffer_count = 64;
-
  public:
   /**
    * Constructor.

intern/ghost/intern/GHOST_ContextCGL.mm

@@ -529,8 +529,7 @@ void GHOST_ContextCGL::metalInit()
     id<MTLDevice> device = m_metalLayer.device;
 
     /* Create a command queue for blit/present operation. */
-    m_metalCmdQueue = (MTLCommandQueue *)[device
+    m_metalCmdQueue = (MTLCommandQueue *)[device newCommandQueue];
-        newCommandQueueWithMaxCommandBufferCount:GHOST_ContextCGL::max_command_buffer_count];
     [m_metalCmdQueue retain];
 
     /* Create shaders for blit operation. */

intern/ghost/intern/GHOST_ContextVK.cpp

@@ -311,12 +311,12 @@ GHOST_TSuccess GHOST_ContextVK::getVulkanBackbuffer(void *image,
 GHOST_TSuccess GHOST_ContextVK::getVulkanHandles(void *r_instance,
                                                  void *r_physical_device,
                                                  void *r_device,
-                                                 uint32_t *r_graphic_queue_family)
+                                                 uint32_t *r_graphic_queue_familly)
 {
   *((VkInstance *)r_instance) = m_instance;
   *((VkPhysicalDevice *)r_physical_device) = m_physical_device;
   *((VkDevice *)r_device) = m_device;
-  *r_graphic_queue_family = m_queue_family_graphic;
+  *r_graphic_queue_familly = m_queue_family_graphic;
 
   return GHOST_kSuccess;
 }
@@ -520,14 +520,13 @@ static GHOST_TSuccess getGraphicQueueFamily(VkPhysicalDevice device, uint32_t *r
 
   *r_queue_index = 0;
   for (const auto &queue_family : queue_families) {
-    if ((queue_family.queueFlags & VK_QUEUE_GRAPHICS_BIT) &&
+    if (queue_family.queueFlags & VK_QUEUE_GRAPHICS_BIT) {
-        (queue_family.queueFlags & VK_QUEUE_COMPUTE_BIT)) {
       return GHOST_kSuccess;
     }
     (*r_queue_index)++;
   }
 
-  fprintf(stderr, "Couldn't find any Graphic queue family on selected device\n");
+  fprintf(stderr, "Couldn't find any Graphic queue familly on selected device\n");
   return GHOST_kFailure;
 }
 
@@ -552,7 +551,7 @@ static GHOST_TSuccess getPresetQueueFamily(VkPhysicalDevice device,
     (*r_queue_index)++;
   }
 
-  fprintf(stderr, "Couldn't find any Present queue family on selected device\n");
+  fprintf(stderr, "Couldn't find any Present queue familly on selected device\n");
   return GHOST_kFailure;
 }
 

intern/ghost/intern/GHOST_ContextVK.h

@@ -113,7 +113,7 @@ class GHOST_ContextVK : public GHOST_Context {
   GHOST_TSuccess getVulkanHandles(void *r_instance,
                                   void *r_physical_device,
                                   void *r_device,
-                                  uint32_t *r_graphic_queue_family);
+                                  uint32_t *r_graphic_queue_familly);
   /**
    * Gets the Vulkan framebuffer related resource handles associated with the Vulkan context.
    * Needs to be called after each swap events as the framebuffer will change.

intern/ghost/intern/GHOST_ContextWGL.cpp

@@ -127,8 +127,7 @@ GHOST_TSuccess GHOST_ContextWGL::releaseDrawingContext()
 /* Ron Fosner's code for weighting pixel formats and forcing software.
  * See http://www.opengl.org/resources/faq/technical/weight.cpp
  */
-ATTR_NO_ASAN static int weight_pixel_format(PIXELFORMATDESCRIPTOR &pfd,
+static int weight_pixel_format(PIXELFORMATDESCRIPTOR &pfd, PIXELFORMATDESCRIPTOR &preferredPFD)
-                                            PIXELFORMATDESCRIPTOR &preferredPFD)
 {
   int weight = 0;
 
@@ -163,7 +162,7 @@ ATTR_NO_ASAN static int weight_pixel_format(PIXELFORMATDESCRIPTOR &pfd,
  * A modification of Ron Fosner's replacement for ChoosePixelFormat
  * returns 0 on error, else returns the pixel format number to be used
  */
-ATTR_NO_ASAN static int choose_pixel_format_legacy(HDC hDC, PIXELFORMATDESCRIPTOR &preferredPFD)
+static int choose_pixel_format_legacy(HDC hDC, PIXELFORMATDESCRIPTOR &preferredPFD)
 {
   int iPixelFormat = 0;
   int weight = 0;
@@ -216,7 +215,7 @@ ATTR_NO_ASAN static int choose_pixel_format_legacy(HDC hDC, PIXELFORMATDESCRIPTO
  * There is no generic way to clone the lpParam parameter,
  * so the caller is responsible for cloning it themselves.
  */
-ATTR_NO_ASAN static HWND clone_window(HWND hWnd, LPVOID lpParam)
+static HWND clone_window(HWND hWnd, LPVOID lpParam)
 {
   int count;
 
@@ -479,7 +478,7 @@ int GHOST_ContextWGL::_choose_pixel_format_arb_1(bool stereoVisual, bool needAlp
   return iPixelFormat;
 }
 
-ATTR_NO_ASAN int GHOST_ContextWGL::choose_pixel_format_arb(bool stereoVisual, bool needAlpha)
+int GHOST_ContextWGL::choose_pixel_format_arb(bool stereoVisual, bool needAlpha)
 {
   int iPixelFormat;
 
@@ -506,7 +505,7 @@ static void reportContextString(const char *name, const char *dummy, const char
 }
 #endif
 
-ATTR_NO_ASAN GHOST_TSuccess GHOST_ContextWGL::initializeDrawingContext()
+GHOST_TSuccess GHOST_ContextWGL::initializeDrawingContext()
 {
   SetLastError(NO_ERROR);
 

intern/ghost/intern/GHOST_SystemWayland.cpp

@@ -82,8 +82,6 @@
 #include "CLG_log.h"
 
 #ifdef USE_EVENT_BACKGROUND_THREAD
-#  include "GHOST_TimerTask.h"
-
 #  include <pthread.h>
 #endif
 
@@ -217,15 +215,13 @@ static bool use_gnome_confine_hack = false;
 /**
  * The event codes are used to differentiate from which mouse button an event comes from.
  */
-enum {
+#define BTN_LEFT 0x110
-  BTN_LEFT = 0x110,
+#define BTN_RIGHT 0x111
-  BTN_RIGHT = 0x111,
+#define BTN_MIDDLE 0x112
-  BTN_MIDDLE = 0x112,
+#define BTN_SIDE 0x113
-  BTN_SIDE = 0x113,
+#define BTN_EXTRA 0x114
-  BTN_EXTRA = 0x114,
+#define BTN_FORWARD 0x115
-  BTN_FORWARD = 0x115,
+#define BTN_BACK 0x116
-  BTN_BACK = 0x116
-};
 // #define BTN_TASK 0x117 /* UNUSED. */
 
 /**
@@ -236,34 +232,28 @@ enum {
  * at the Blender studio, having the button closest to the nib be MMB is preferable,
  * so use this as a default. If needs be - swapping these could be a preference.
  */
-enum {
+#define BTN_STYLUS 0x14b  /* Use as middle-mouse. */
-  /** Use as middle-mouse. */
+#define BTN_STYLUS2 0x14c /* Use as right-mouse. */
-  BTN_STYLUS = 0x14b,
+/* NOTE(@campbellbarton): Map to an additional button (not sure which hardware uses this). */
-  /** Use as right-mouse. */
+#define BTN_STYLUS3 0x149
-  BTN_STYLUS2 = 0x14c,
-  /** NOTE(@campbellbarton): Map to an additional button (not sure which hardware uses this). */
-  BTN_STYLUS3 = 0x149,
-};
 
 /**
  * Keyboard scan-codes.
  */
-enum {
+#define KEY_GRAVE 41
-  KEY_GRAVE = 41,
 
 #ifdef USE_NON_LATIN_KB_WORKAROUND
-  KEY_1 = 2,
+#  define KEY_1 2
-  KEY_2 = 3,
+#  define KEY_2 3
-  KEY_3 = 4,
+#  define KEY_3 4
-  KEY_4 = 5,
+#  define KEY_4 5
-  KEY_5 = 6,
+#  define KEY_5 6
-  KEY_6 = 7,
+#  define KEY_6 7
-  KEY_7 = 8,
+#  define KEY_7 8
-  KEY_8 = 9,
+#  define KEY_8 9
-  KEY_9 = 10,
+#  define KEY_9 10
-  KEY_0 = 11,
+#  define KEY_0 11
 #endif
-};
 
 /** \} */
 
@@ -770,12 +760,7 @@ struct GWL_Seat {
     int32_t rate = 0;
     /** Time (milliseconds) after which to start repeating keys. */
     int32_t delay = 0;
-    /**
+    /** Timer for key repeats. */
-     * Timer for key repeats.
-     *
-     * \note For as long as #USE_EVENT_BACKGROUND_THREAD is defined, any access to this
-     * (including null checks, must lock `timer_mutex` first.
-     */
     GHOST_ITimerTask *timer = nullptr;
   } key_repeat;
 
@@ -839,42 +824,6 @@ static bool gwl_seat_key_depressed_suppress_warning(const GWL_Seat *seat)
   return suppress_warning;
 }
 
-/**
- * \note Caller must lock `timer_mutex`.
- */
-static void gwl_seat_key_repeat_timer_add(GWL_Seat *seat,
-                                          GHOST_TimerProcPtr key_repeat_fn,
-                                          GHOST_TUserDataPtr payload,
-                                          const bool use_delay)
-{
-  GHOST_SystemWayland *system = seat->system;
-  const uint64_t time_step = 1000 / seat->key_repeat.rate;
-  const uint64_t time_start = use_delay ? seat->key_repeat.delay : time_step;
-#ifdef USE_EVENT_BACKGROUND_THREAD
-  GHOST_TimerTask *timer = new GHOST_TimerTask(
-      system->getMilliSeconds() + time_start, time_step, key_repeat_fn, payload);
-  seat->key_repeat.timer = timer;
-  system->ghost_timer_manager()->addTimer(timer);
-#else
-  seat->key_repeat.timer = system->installTimer(time_start, time_step, key_repeat_fn, payload);
-#endif
-}
-
-/**
- * \note The caller must lock `timer_mutex`.
- */
-static void gwl_seat_key_repeat_timer_remove(GWL_Seat *seat)
-{
-  GHOST_SystemWayland *system = seat->system;
-#ifdef USE_EVENT_BACKGROUND_THREAD
-  system->ghost_timer_manager()->removeTimer(
-      static_cast<GHOST_TimerTask *>(seat->key_repeat.timer));
-#else
-  system->removeTimer(seat->key_repeat.timer);
-#endif
-  seat->key_repeat.timer = nullptr;
-}
-
 /** \} */
 
 /* -------------------------------------------------------------------- */
@@ -949,16 +898,6 @@ struct GWL_Display {
   /** Guard against multiple threads accessing `events_pending` at once. */
   std::mutex events_pending_mutex;
 
-  /**
-   * A separate timer queue, needed so the WAYLAND thread can lock access.
-   * Using the system's #GHOST_Sysem::getTimerManager is not thread safe because
-   * access to the timer outside of WAYLAND specific logic will not lock.
-   *
-   * Needed because #GHOST_System::dispatchEvents fires timers
-   * outside of WAYLAND (without locking the `timer_mutex`).
-   */
-  GHOST_TimerManager *ghost_timer_manager;
-
 #endif /* USE_EVENT_BACKGROUND_THREAD */
 };
 
@@ -975,9 +914,6 @@ static void gwl_display_destroy(GWL_Display *display)
     ghost_wl_display_lock_without_input(display->wl_display, display->system->server_mutex);
     display->events_pthread_is_active = false;
   }
-
-  delete display->ghost_timer_manager;
-  display->ghost_timer_manager = nullptr;
 #endif
 
   /* For typical WAYLAND use this will always be set.
@@ -3774,15 +3710,10 @@ static void keyboard_handle_leave(void *data,
   GWL_Seat *seat = static_cast<GWL_Seat *>(data);
   seat->keyboard.wl_surface_window = nullptr;
 
-  {
-#ifdef USE_EVENT_BACKGROUND_THREAD
-    std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
-#endif
   /* Losing focus must stop repeating text. */
   if (seat->key_repeat.timer) {
     keyboard_handle_key_repeat_cancel(seat);
   }
-  }
 
 #ifdef USE_GNOME_KEYBOARD_SUPPRESS_WARNING
   seat->key_depressed_suppress_warning.any_mod_held = false;
@@ -3841,32 +3772,36 @@ static xkb_keysym_t xkb_state_key_get_one_sym_without_modifiers(
   return sym;
 }
 
-/**
- * \note Caller must lock `timer_mutex`.
- */
 static void keyboard_handle_key_repeat_cancel(GWL_Seat *seat)
 {
+#ifdef USE_EVENT_BACKGROUND_THREAD
+  std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
+#endif
   GHOST_ASSERT(seat->key_repeat.timer != nullptr, "Caller much check for timer");
   delete static_cast<GWL_KeyRepeatPlayload *>(seat->key_repeat.timer->getUserData());
-
+  seat->system->removeTimer(seat->key_repeat.timer);
-  gwl_seat_key_repeat_timer_remove(seat);
+  seat->key_repeat.timer = nullptr;
 }
 
 /**
  * Restart the key-repeat timer.
  * \param use_delay: When false, use the interval
  * (prevents pause when the setting changes while the key is held).
- *
- * \note Caller must lock `timer_mutex`.
  */
 static void keyboard_handle_key_repeat_reset(GWL_Seat *seat, const bool use_delay)
 {
+#ifdef USE_EVENT_BACKGROUND_THREAD
+  std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
+#endif
   GHOST_ASSERT(seat->key_repeat.timer != nullptr, "Caller much check for timer");
-  GHOST_TimerProcPtr key_repeat_fn = seat->key_repeat.timer->getTimerProc();
+  GHOST_SystemWayland *system = seat->system;
+  GHOST_ITimerTask *timer = seat->key_repeat.timer;
+  GHOST_TimerProcPtr key_repeat_fn = timer->getTimerProc();
   GHOST_TUserDataPtr payload = seat->key_repeat.timer->getUserData();
-
+  seat->system->removeTimer(seat->key_repeat.timer);
-  gwl_seat_key_repeat_timer_remove(seat);
+  const uint64_t time_step = 1000 / seat->key_repeat.rate;
-  gwl_seat_key_repeat_timer_add(seat, key_repeat_fn, payload, use_delay);
+  const uint64_t time_start = use_delay ? seat->key_repeat.delay : time_step;
+  seat->key_repeat.timer = system->installTimer(time_start, time_step, key_repeat_fn, payload);
 }
 
 static void keyboard_handle_key(void *data,
@@ -3905,11 +3840,6 @@ static void keyboard_handle_key(void *data,
       break;
   }
 
-#ifdef USE_EVENT_BACKGROUND_THREAD
-  /* Any access to `seat->key_repeat.timer` must lock. */
-  std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
-#endif
-
   struct GWL_KeyRepeatPlayload *key_repeat_payload = nullptr;
 
   /* Delete previous timer. */
@@ -3948,14 +3878,23 @@ static void keyboard_handle_key(void *data,
         break;
       }
       case RESET: {
+#ifdef USE_EVENT_BACKGROUND_THREAD
+        std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
+#endif
         /* The payload will be added again. */
-        gwl_seat_key_repeat_timer_remove(seat);
+        seat->system->removeTimer(seat->key_repeat.timer);
+        seat->key_repeat.timer = nullptr;
         break;
       }
       case CANCEL: {
+#ifdef USE_EVENT_BACKGROUND_THREAD
+        std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
+#endif
         delete key_repeat_payload;
         key_repeat_payload = nullptr;
-        gwl_seat_key_repeat_timer_remove(seat);
+
+        seat->system->removeTimer(seat->key_repeat.timer);
+        seat->key_repeat.timer = nullptr;
         break;
       }
     }
@@ -4009,8 +3948,8 @@ static void keyboard_handle_key(void *data,
                                                            utf8_buf));
       }
     };
-
+    seat->key_repeat.timer = seat->system->installTimer(
-    gwl_seat_key_repeat_timer_add(seat, key_repeat_fn, key_repeat_payload, true);
+        seat->key_repeat.delay, 1000 / seat->key_repeat.rate, key_repeat_fn, key_repeat_payload);
   }
 }
 
@@ -4035,14 +3974,9 @@ static void keyboard_handle_modifiers(void *data,
 
   /* A modifier changed so reset the timer,
    * see comment in #keyboard_handle_key regarding this behavior. */
-  {
-#ifdef USE_EVENT_BACKGROUND_THREAD
-    std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
-#endif
   if (seat->key_repeat.timer) {
     keyboard_handle_key_repeat_reset(seat, true);
   }
-  }
 
 #ifdef USE_GNOME_KEYBOARD_SUPPRESS_WARNING
   seat->key_depressed_suppress_warning.any_mod_held = mods_depressed != 0;
@@ -4060,15 +3994,10 @@ static void keyboard_repeat_handle_info(void *data,
   seat->key_repeat.rate = rate;
   seat->key_repeat.delay = delay;
 
-  {
-#ifdef USE_EVENT_BACKGROUND_THREAD
-    std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
-#endif
   /* Unlikely possible this setting changes while repeating. */
   if (seat->key_repeat.timer) {
     keyboard_handle_key_repeat_reset(seat, false);
   }
-  }
 }
 
 static const struct wl_keyboard_listener keyboard_listener = {
@@ -4338,15 +4267,9 @@ static void gwl_seat_capability_keyboard_disable(GWL_Seat *seat)
   if (!seat->wl_keyboard) {
     return;
   }
-
-  {
-#ifdef USE_EVENT_BACKGROUND_THREAD
-    std::lock_guard lock_timer_guard{*seat->system->timer_mutex};
-#endif
   if (seat->key_repeat.timer) {
     keyboard_handle_key_repeat_cancel(seat);
   }
-  }
   wl_keyboard_destroy(seat->wl_keyboard);
   seat->wl_keyboard = nullptr;
 }
@@ -5480,8 +5403,6 @@ GHOST_SystemWayland::GHOST_SystemWayland(bool background)
 
 #ifdef USE_EVENT_BACKGROUND_THREAD
   gwl_display_event_thread_create(display_);
-
-  display_->ghost_timer_manager = new GHOST_TimerManager();
 #endif
 }
 
@@ -5562,16 +5483,10 @@ bool GHOST_SystemWayland::processEvents(bool waitForEvent)
 #endif /* USE_EVENT_BACKGROUND_THREAD */
 
   {
-    const uint64_t now = getMilliSeconds();
 #ifdef USE_EVENT_BACKGROUND_THREAD
-    {
     std::lock_guard lock_timer_guard{*display_->system->timer_mutex};
-      if (ghost_timer_manager()->fireTimers(now)) {
-        any_processed = true;
-      }
-    }
 #endif
-    if (getTimerManager()->fireTimers(now)) {
+    if (getTimerManager()->fireTimers(getMilliSeconds())) {
       any_processed = true;
     }
   }
@@ -6794,13 +6709,6 @@ struct wl_shm *GHOST_SystemWayland::wl_shm() const
   return display_->wl_shm;
 }
 
-#ifdef USE_EVENT_BACKGROUND_THREAD
-GHOST_TimerManager *GHOST_SystemWayland::ghost_timer_manager()
-{
-  return display_->ghost_timer_manager;
-}
-#endif
-
 /** \} */
 
 /* -------------------------------------------------------------------- */

intern/ghost/intern/GHOST_SystemWayland.h

@@ -165,16 +165,6 @@ class GHOST_SystemWayland : public GHOST_System {
 
   bool cursor_grab_use_software_display_get(const GHOST_TGrabCursorMode mode);
 
-#ifdef USE_EVENT_BACKGROUND_THREAD
-  /**
-   * Return a separate WAYLAND local timer manager to #GHOST_System::getTimerManager
-   * Manipulation & access must lock with #GHOST_WaylandSystem::server_mutex.
-   *
-   * See #GWL_Display::ghost_timer_manager doc-string for details on why this is needed.
-   */
-  GHOST_TimerManager *ghost_timer_manager();
-#endif
-
   /* WAYLAND direct-data access. */
 
   struct wl_display *wl_display();
@@ -243,14 +233,7 @@ class GHOST_SystemWayland : public GHOST_System {
    * from running at the same time. */
   std::mutex *server_mutex = nullptr;
 
-  /**
+  /** Threads must lock this before manipulating timers. */
-   * Threads must lock this before manipulating #GWL_Display::ghost_timer_manager.
-   *
-   * \note Using a separate lock to `server_mutex` is necessary because the
-   * server lock is already held when calling `ghost_wl_display_event_pump`.
-   * If manipulating the timer used the `server_mutex`, event pump can indirectly
-   * handle key up/down events which would lock `server_mutex` causing a dead-lock.
-   */
   std::mutex *timer_mutex = nullptr;
 
   std::thread::id main_thread_id;

intern/ghost/intern/GHOST_WindowWin32.cpp

@@ -39,7 +39,7 @@ extern "C" {
 __declspec(dllexport) DWORD NvOptimusEnablement = 0x00000001;
 }
 
-ATTR_NO_ASAN GHOST_WindowWin32::GHOST_WindowWin32(GHOST_SystemWin32 *system,
+GHOST_WindowWin32::GHOST_WindowWin32(GHOST_SystemWin32 *system,
                                      const char *title,
                                      int32_t left,
                                      int32_t top,
@@ -579,7 +579,7 @@ GHOST_TSuccess GHOST_WindowWin32::invalidate()
   return success;
 }
 
-ATTR_NO_ASAN GHOST_Context *GHOST_WindowWin32::newDrawingContext(GHOST_TDrawingContextType type)
+GHOST_Context *GHOST_WindowWin32::newDrawingContext(GHOST_TDrawingContextType type)
 {
   if (type == GHOST_kDrawingContextTypeOpenGL) {
     GHOST_Context *context;

intern/ghost/intern/GHOST_Wintab.cpp

@@ -7,20 +7,12 @@
 #define _USE_MATH_DEFINES
 
 #include "GHOST_Wintab.h"
-#include <cstdio>
-
-static void wintab_load_error(const char *func)
-{
-  fprintf(stderr, "corrupted wintab32.dll; missing %s\n", func);
-}
 
 GHOST_Wintab *GHOST_Wintab::loadWintab(HWND hwnd)
 {
   /* Load Wintab library if available. */
   auto handle = unique_hmodule(::LoadLibrary("Wintab32.dll"), &::FreeLibrary);
   if (!handle) {
-    fprintf(stderr, "Could not find wintab32.dll\n");
-
     return nullptr;
   }
 
@@ -28,61 +20,51 @@ GHOST_Wintab *GHOST_Wintab::loadWintab(HWND hwnd)
 
   auto info = (GHOST_WIN32_WTInfo)::GetProcAddress(handle.get(), "WTInfoA");
   if (!info) {
-    wintab_load_error("WTInfoA");
     return nullptr;
   }
 
   auto open = (GHOST_WIN32_WTOpen)::GetProcAddress(handle.get(), "WTOpenA");
   if (!open) {
-    wintab_load_error("WTOpenA");
     return nullptr;
   }
 
   auto get = (GHOST_WIN32_WTGet)::GetProcAddress(handle.get(), "WTGetA");
   if (!get) {
-    wintab_load_error("WTGetA");
     return nullptr;
   }
 
   auto set = (GHOST_WIN32_WTSet)::GetProcAddress(handle.get(), "WTSetA");
   if (!set) {
-    wintab_load_error("WTSetA");
     return nullptr;
   }
 
   auto close = (GHOST_WIN32_WTClose)::GetProcAddress(handle.get(), "WTClose");
   if (!close) {
-    wintab_load_error("WTClose");
     return nullptr;
   }
 
   auto packetsGet = (GHOST_WIN32_WTPacketsGet)::GetProcAddress(handle.get(), "WTPacketsGet");
   if (!packetsGet) {
-    wintab_load_error("WTPacketGet");
     return nullptr;
   }
 
   auto queueSizeGet = (GHOST_WIN32_WTQueueSizeGet)::GetProcAddress(handle.get(), "WTQueueSizeGet");
   if (!queueSizeGet) {
-    wintab_load_error("WTQueueSizeGet");
     return nullptr;
   }
 
   auto queueSizeSet = (GHOST_WIN32_WTQueueSizeSet)::GetProcAddress(handle.get(), "WTQueueSizeSet");
   if (!queueSizeSet) {
-    wintab_load_error("WTQueueSizeSet");
     return nullptr;
   }
 
   auto enable = (GHOST_WIN32_WTEnable)::GetProcAddress(handle.get(), "WTEnable");
   if (!enable) {
-    wintab_load_error("WTEnable");
     return nullptr;
   }
 
   auto overlap = (GHOST_WIN32_WTOverlap)::GetProcAddress(handle.get(), "WTOverlap");
   if (!overlap) {
-    wintab_load_error("WTOverlap");
     return nullptr;
   }
 
@@ -90,8 +72,6 @@ GHOST_Wintab *GHOST_Wintab::loadWintab(HWND hwnd)
 
   LOGCONTEXT lc = {0};
   if (!info(WTI_DEFSYSCTX, 0, &lc)) {
-    fprintf(stderr, "Failed to initialize Wintab driver\n");
-
     return nullptr;
   }
 
@@ -102,8 +82,6 @@ GHOST_Wintab *GHOST_Wintab::loadWintab(HWND hwnd)
   /* The Wintab spec says we must open the context disabled if we are using cursor masks. */
   auto hctx = unique_hctx(open(hwnd, &lc, FALSE), close);
   if (!hctx) {
-    fprintf(stderr, "Failed to open Wintab driver\n");
-
     return nullptr;
   }
 
@@ -210,7 +188,7 @@ GHOST_Wintab::GHOST_Wintab(unique_hmodule handle,
       m_context{std::move(hctx)},
       m_tabletCoord{tablet},
       m_systemCoord{system},
-      m_pkts{(size_t)queueSize}
+      m_pkts{queueSize}
 {
   m_fpInfo(WTI_INTERFACE, IFC_NDEVICES, &m_numDevices);
   WINTAB_PRINTF("Wintab Devices: %d\n", m_numDevices);

intern/guardedalloc/CMakeLists.txt

@@ -9,7 +9,6 @@ set(INC
   .
   ..
   ../atomic
-  ../../source/blender/blenlib
 )
 
 set(INC_SYS

intern/guardedalloc/MEM_guardedalloc.h

@@ -66,12 +66,6 @@ extern short (*MEM_testN)(void *vmemh);
  * NULL-safe; will return NULL when receiving a NULL pointer. */
 extern void *(*MEM_dupallocN)(const void *vmemh) /* ATTR_MALLOC */ ATTR_WARN_UNUSED_RESULT;
 
-/**
- * Duplicates a block of memory, and returns a pointer to the
- * newly allocated block. */
-extern void *(*MEM_dupallocN_id)(const void *vmemh,
-                                 const char *str) /* ATTR_MALLOC */ ATTR_WARN_UNUSED_RESULT;
-
 /**
  * Reallocates a block of memory, and returns pointer to the newly
  * allocated block, the old one is freed. this is not as optimized
@@ -253,8 +247,6 @@ void MEM_use_lockfree_allocator(void);
  * NOTE: The switch between allocator types can only happen before any allocation did happen. */
 void MEM_use_guarded_allocator(void);
 
-#define MEM_dupallocN(vmemh) MEM_dupallocN_id(vmemh, __func__)
-
 #ifdef __cplusplus
 }
 #endif /* __cplusplus */

intern/guardedalloc/intern/mallocn.c

@@ -26,7 +26,6 @@ const char *malloc_conf = "background_thread:true,dirty_decay_ms:4000";
 size_t (*MEM_allocN_len)(const void *vmemh) = MEM_lockfree_allocN_len;
 void (*MEM_freeN)(void *vmemh) = MEM_lockfree_freeN;
 void *(*MEM_dupallocN)(const void *vmemh) = MEM_lockfree_dupallocN;
-void *(*MEM_dupallocN_id)(const void *vmemh, const char *str) = MEM_lockfree_dupallocN_id;
 void *(*MEM_reallocN_id)(void *vmemh, size_t len, const char *str) = MEM_lockfree_reallocN_id;
 void *(*MEM_recallocN_id)(void *vmemh, size_t len, const char *str) = MEM_lockfree_recallocN_id;
 void *(*MEM_callocN)(size_t len, const char *str) = MEM_lockfree_callocN;
@@ -109,7 +108,6 @@ void MEM_use_lockfree_allocator(void)
   MEM_allocN_len = MEM_lockfree_allocN_len;
   MEM_freeN = MEM_lockfree_freeN;
   MEM_dupallocN = MEM_lockfree_dupallocN;
-  MEM_dupallocN_id = MEM_lockfree_dupallocN_id;
   MEM_reallocN_id = MEM_lockfree_reallocN_id;
   MEM_recallocN_id = MEM_lockfree_recallocN_id;
   MEM_callocN = MEM_lockfree_callocN;
@@ -142,7 +140,6 @@ void MEM_use_guarded_allocator(void)
   MEM_allocN_len = MEM_guarded_allocN_len;
   MEM_freeN = MEM_guarded_freeN;
   MEM_dupallocN = MEM_guarded_dupallocN;
-  MEM_dupallocN_id = MEM_guarded_dupallocN_id;
   MEM_reallocN_id = MEM_guarded_reallocN_id;
   MEM_recallocN_id = MEM_guarded_recallocN_id;
   MEM_callocN = MEM_guarded_callocN;

intern/guardedalloc/intern/mallocn_guarded_impl.c

@@ -287,31 +287,6 @@ void *MEM_guarded_dupallocN(const void *vmemh)
   return newp;
 }
 
-void *MEM_guarded_dupallocN_id(const void *vmemh, const char *str)
-{
-  void *newp = NULL;
-
-  if (vmemh) {
-    const MemHead *memh = vmemh;
-    memh--;
-
-    if (LIKELY(memh->alignment == 0)) {
-      newp = MEM_guarded_mallocN(memh->len, str);
-    }
-    else {
-      newp = MEM_guarded_mallocN_aligned(memh->len, (size_t)memh->alignment, str);
-    }
-
-    if (newp == NULL) {
-      return NULL;
-    }
-
-    memcpy(newp, vmemh, memh->len);
-  }
-
-  return newp;
-}
-
 void *MEM_guarded_reallocN_id(void *vmemh, size_t len, const char *str)
 {
   void *newp = NULL;

intern/guardedalloc/intern/mallocn_intern.h

@@ -101,8 +101,6 @@ void memory_usage_peak_reset(void);
 size_t MEM_lockfree_allocN_len(const void *vmemh) ATTR_WARN_UNUSED_RESULT;
 void MEM_lockfree_freeN(void *vmemh);
 void *MEM_lockfree_dupallocN(const void *vmemh) ATTR_MALLOC ATTR_WARN_UNUSED_RESULT;
-void *MEM_lockfree_dupallocN_id(const void *vmemh,
-                                const char *str) ATTR_MALLOC ATTR_WARN_UNUSED_RESULT;
 void *MEM_lockfree_reallocN_id(void *vmemh,
                                size_t len,
                                const char *str) ATTR_MALLOC ATTR_WARN_UNUSED_RESULT
@@ -147,8 +145,6 @@ void MEM_lockfree_name_ptr_set(void *vmemh, const char *str);
 size_t MEM_guarded_allocN_len(const void *vmemh) ATTR_WARN_UNUSED_RESULT;
 void MEM_guarded_freeN(void *vmemh);
 void *MEM_guarded_dupallocN(const void *vmemh) ATTR_MALLOC ATTR_WARN_UNUSED_RESULT;
-void *MEM_guarded_dupallocN_id(const void *vmemh,
-                               const char *str) ATTR_MALLOC ATTR_WARN_UNUSED_RESULT;
 void *MEM_guarded_reallocN_id(void *vmemh,
                               size_t len,
                               const char *str) ATTR_MALLOC ATTR_WARN_UNUSED_RESULT

intern/guardedalloc/intern/mallocn_lockfree_impl.c

@@ -15,7 +15,6 @@
 #include "MEM_guardedalloc.h"
 
 /* to ensure strict conversions */
-#include "../../source/blender/blenlib/BLI_asan.h"
 #include "../../source/blender/blenlib/BLI_strict_flags.h"
 
 #include "atomic_ops.h"
@@ -45,21 +44,6 @@ enum {
 #define MEMHEAD_IS_ALIGNED(memhead) ((memhead)->len & (size_t)MEMHEAD_ALIGN_FLAG)
 #define MEMHEAD_LEN(memhead) ((memhead)->len & ~((size_t)(MEMHEAD_ALIGN_FLAG)))
 
-#define MEM_POISON_MEMHEAD(vmemh) BLI_asan_poison(MEMHEAD_FROM_PTR(vmemh), sizeof(MemHead))
-#define MEM_UNPOISON_MEMHEAD(vmemh) BLI_asan_unpoison(MEMHEAD_FROM_PTR(vmemh), sizeof(MemHead))
-
-/* Uncomment this to have proper peak counter. */
-#define USE_ATOMIC_MAX
-
-MEM_INLINE void update_maximum(size_t *maximum_value, size_t value)
-{
-#ifdef USE_ATOMIC_MAX
-  atomic_fetch_and_update_max_z(maximum_value, value);
-#else
-  *maximum_value = value > *maximum_value ? value : *maximum_value;
-#endif
-}
-
 #ifdef __GNUC__
 __attribute__((format(printf, 1, 2)))
 #endif
@@ -82,13 +66,7 @@ print_error(const char *str, ...)
 size_t MEM_lockfree_allocN_len(const void *vmemh)
 {
   if (LIKELY(vmemh)) {
-    size_t ret;
+    return MEMHEAD_LEN(MEMHEAD_FROM_PTR(vmemh));
-
-    MEM_UNPOISON_MEMHEAD(vmemh);
-    ret = MEMHEAD_FROM_PTR(vmemh)->len & ~((size_t)(MEMHEAD_ALIGN_FLAG));
-    MEM_POISON_MEMHEAD(vmemh);
-
-    return ret;
   }
 
   return 0;
@@ -109,8 +87,6 @@ void MEM_lockfree_freeN(void *vmemh)
   }
 
   MemHead *memh = MEMHEAD_FROM_PTR(vmemh);
-  MEM_UNPOISON_MEMHEAD(vmemh);
-
   size_t len = MEMHEAD_LEN(memh);
 
   memory_usage_block_free(len);
@@ -133,9 +109,6 @@ void *MEM_lockfree_dupallocN(const void *vmemh)
   if (vmemh) {
     MemHead *memh = MEMHEAD_FROM_PTR(vmemh);
     const size_t prev_size = MEM_lockfree_allocN_len(vmemh);
-
-    MEM_UNPOISON_MEMHEAD(vmemh);
-
     if (UNLIKELY(MEMHEAD_IS_ALIGNED(memh))) {
       MemHeadAligned *memh_aligned = MEMHEAD_ALIGNED_FROM_PTR(vmemh);
       newp = MEM_lockfree_mallocN_aligned(
@@ -144,31 +117,6 @@ void *MEM_lockfree_dupallocN(const void *vmemh)
     else {
       newp = MEM_lockfree_mallocN(prev_size, "dupli_malloc");
     }
-
-    MEM_POISON_MEMHEAD(vmemh);
-    memcpy(newp, vmemh, prev_size);
-  }
-  return newp;
-}
-
-void *MEM_lockfree_dupallocN_id(const void *vmemh, const char *str)
-{
-  void *newp = NULL;
-  if (vmemh) {
-    MemHead *memh = MEMHEAD_FROM_PTR(vmemh);
-    const size_t prev_size = MEM_lockfree_allocN_len(vmemh);
-
-    MEM_UNPOISON_MEMHEAD(vmemh);
-
-    if (UNLIKELY(MEMHEAD_IS_ALIGNED(memh))) {
-      MemHeadAligned *memh_aligned = MEMHEAD_ALIGNED_FROM_PTR(vmemh);
-      newp = MEM_lockfree_mallocN_aligned(prev_size, (size_t)memh_aligned->alignment, str);
-    }
-    else {
-      newp = MEM_lockfree_mallocN(prev_size, str);
-    }
-
-    MEM_POISON_MEMHEAD(vmemh);
     memcpy(newp, vmemh, prev_size);
   }
   return newp;
@@ -182,8 +130,6 @@ void *MEM_lockfree_reallocN_id(void *vmemh, size_t len, const char *str)
     MemHead *memh = MEMHEAD_FROM_PTR(vmemh);
     size_t old_len = MEM_lockfree_allocN_len(vmemh);
 
-    MEM_UNPOISON_MEMHEAD(vmemh);
-
     if (LIKELY(!MEMHEAD_IS_ALIGNED(memh))) {
       newp = MEM_lockfree_mallocN(len, "realloc");
     }
@@ -192,8 +138,6 @@ void *MEM_lockfree_reallocN_id(void *vmemh, size_t len, const char *str)
       newp = MEM_lockfree_mallocN_aligned(len, (size_t)memh_aligned->alignment, "realloc");
     }
 
-    MEM_POISON_MEMHEAD(vmemh);
-
     if (newp) {
       if (len < old_len) {
         /* shrink */
@@ -222,8 +166,6 @@ void *MEM_lockfree_recallocN_id(void *vmemh, size_t len, const char *str)
     MemHead *memh = MEMHEAD_FROM_PTR(vmemh);
     size_t old_len = MEM_lockfree_allocN_len(vmemh);
 
-    MEM_UNPOISON_MEMHEAD(vmemh);
-
     if (LIKELY(!MEMHEAD_IS_ALIGNED(memh))) {
       newp = MEM_lockfree_mallocN(len, "recalloc");
     }
@@ -231,7 +173,6 @@ void *MEM_lockfree_recallocN_id(void *vmemh, size_t len, const char *str)
       MemHeadAligned *memh_aligned = MEMHEAD_ALIGNED_FROM_PTR(vmemh);
       newp = MEM_lockfree_mallocN_aligned(len, (size_t)memh_aligned->alignment, "recalloc");
     }
-    MEM_POISON_MEMHEAD(vmemh);
 
     if (newp) {
       if (len < old_len) {
@@ -270,7 +211,6 @@ void *MEM_lockfree_callocN(size_t len, const char *str)
     memh->len = len;
     memory_usage_block_alloc(len);
 
-    MEM_POISON_MEMHEAD(PTR_FROM_MEMHEAD(memh));
     return PTR_FROM_MEMHEAD(memh);
   }
   print_error("Calloc returns null: len=" SIZET_FORMAT " in %s, total %u\n",
@@ -314,8 +254,6 @@ void *MEM_lockfree_mallocN(size_t len, const char *str)
     memh->len = len;
     memory_usage_block_alloc(len);
 
-    MEM_POISON_MEMHEAD(PTR_FROM_MEMHEAD(memh));
-
     return PTR_FROM_MEMHEAD(memh);
   }
   print_error("Malloc returns null: len=" SIZET_FORMAT " in %s, total %u\n",
@@ -385,8 +323,6 @@ void *MEM_lockfree_mallocN_aligned(size_t len, size_t alignment, const char *str
     memh->alignment = (short)alignment;
     memory_usage_block_alloc(len);
 
-    MEM_POISON_MEMHEAD(PTR_FROM_MEMHEAD(memh));
-
     return PTR_FROM_MEMHEAD(memh);
   }
   print_error("Malloc returns null: len=" SIZET_FORMAT " in %s, total %u\n",

intern/instant-meshes/CMakeLists.txt

@@ -1,98 +0,0 @@
-# ***** BEGIN GPL LICENSE BLOCK *****
-#
-# This program is free software; you can redistribute it and/or
-# modify it under the terms of the GNU General Public License
-# as published by the Free Software Foundation; either version 2
-# of the License, or (at your option) any later version.
-#
-# This program is distributed in the hope that it will be useful,
-# but WITHOUT ANY WARRANTY; without even the implied warranty of
-# MERCHANTABILITY or FITNESS FOR A PARTICULAR PURPOSE.  See the
-# GNU General Public License for more details.
-#
-# You should have received a copy of the GNU General Public License
-# along with this program; if not, write to the Free Software Foundation,
-# Inc., 51 Franklin Street, Fifth Floor, Boston, MA 02110-1301, USA.
-#
-# The Original Code is Copyright (C) 2006, Blender Foundation
-# All rights reserved.
-# ***** END GPL LICENSE BLOCK *****
-
-set(INC
-  .
-  ../atomic
-  ../eigen
-  ../guardedalloc
-  ../../extern/Eigen3
-  ext/half
-  ext/dset
-  ext/pss
-  ext/pcg32
-  ../
-  
-)
-
-set(INC_SYS
-  ${GMP_INCLUDE_DIRS}
-)
-
-set(SRC
-  src/aabb.h
-  src/adjacency.cpp
-  src/adjacency.h
-  src/batch.cpp
-  src/batch.h
-  src/bvh.cpp
-  src/bvh.h
-  src/cleanup.cpp
-  src/cleanup.h
-  src/common.h
-  src/dedge.cpp
-  src/dedge.h
-  src/diff.cpp
-  src/extract.cpp
-  src/extract.h
-  src/field.cpp
-  src/field.h
-  src/meshstats.cpp
-  src/meshstats.h
-  src/normal.cpp
-  src/normal.h
-  src/hierarchy.cpp
-  src/hierarchy.h
-  src/reorder.cpp
-  src/reorder.h
-  src/subdivide.cpp
-  src/subdivide.h
-  src/smoothcurve.cpp
-  src/smoothcurve.h
-  src/c_api.cpp
-  instant_meshes_c_api.h
-)
-
-set(LIB
-)
-
-if(WITH_TBB)
-  add_definitions(-DWITH_TBB)
-
-  list(APPEND INC_SYS
-    ${TBB_INCLUDE_DIRS}
-  )
-
-  list(APPEND LIB
-    ${TBB_LIBRARIES}
-  )
-endif()
-
-if(WIN32 AND NOT UNIX)
-  list(APPEND INC_SYS
-    ${PTHREADS_INC}
-  )
-
-  list(APPEND LIB
-    ${PTHREADS_LIBRARIES}
-  )
-endif()
-
-blender_add_lib(bf_intern_instant_meshes "${SRC}" "${INC}" "${INC_SYS}" "${LIB}")

intern/instant-meshes/LICENSE.txt

@@ -1,37 +0,0 @@
-Copyright (c) 2015 Wenzel Jakob, Daniele Panozzo, Marco Tarini,
-and Olga Sorkine-Hornung. All rights reserved.
-
-Redistribution and use in source and binary forms, with or without
-modification, are permitted provided that the following conditions are met:
-
-1. Redistributions of source code must retain the above copyright notice, this
-   list of conditions and the following disclaimer.
-
-2. Redistributions in binary form must reproduce the above copyright notice,
-   this list of conditions and the following disclaimer in the documentation
-   and/or other materials provided with the distribution.
-
-3. Neither the name of the copyright holder nor the names of its contributors
-   may be used to endorse or promote products derived from this software
-   without specific prior written permission.
-
-THIS SOFTWARE IS PROVIDED BY THE COPYRIGHT HOLDERS AND CONTRIBUTORS "AS IS" AND
-ANY EXPRESS OR IMPLIED WARRANTIES, INCLUDING, BUT NOT LIMITED TO, THE IMPLIED
-WARRANTIES OF MERCHANTABILITY AND FITNESS FOR A PARTICULAR PURPOSE ARE
-DISCLAIMED. IN NO EVENT SHALL THE COPYRIGHT HOLDER OR CONTRIBUTORS BE LIABLE
-FOR ANY DIRECT, INDIRECT, INCIDENTAL, SPECIAL, EXEMPLARY, OR CONSEQUENTIAL
-DAMAGES (INCLUDING, BUT NOT LIMITED TO, PROCUREMENT OF SUBSTITUTE GOODS OR
-SERVICES; LOSS OF USE, DATA, OR PROFITS; OR BUSINESS INTERRUPTION) HOWEVER
-CAUSED AND ON ANY THEORY OF LIABILITY, WHETHER IN CONTRACT, STRICT LIABILITY,
-OR TORT (INCLUDING NEGLIGENCE OR OTHERWISE) ARISING IN ANY WAY OUT OF THE USE
-OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
-
-You are under no obligation whatsoever to provide any bug fixes, patches, or
-upgrades to the features, functionality or performance of the source code
-("Enhancements") to anyone; however, if you choose to make your Enhancements
-available either publicly, or directly to the authors of this software, without
-imposing a separate written license agreement for such Enhancements, then you
-hereby grant the following license: a non-exclusive, royalty-free perpetual
-license to install, use, modify, prepare derivative works, incorporate into
-other computer software, distribute, and sublicense such enhancements or
-derivative works thereof, in binary and source code form.

intern/instant-meshes/README.md

@@ -1,72 +0,0 @@
-# Instant Meshes
-[![Build Status](https://travis-ci.org/wjakob/instant-meshes.svg?branch=master)](https://travis-ci.org/wjakob/instant-meshes)
-[![Build status](https://ci.appveyor.com/api/projects/status/dm4kqxhin5uxiey0/branch/master?svg=true)](https://ci.appveyor.com/project/wjakob/instant-meshes/branch/master)
-
-<img width="170" height="166" src="https://github.com/wjakob/instant-meshes/raw/master/resources/icon.png">
-
-This repository contains the interactive meshing software developed as part of the publication
-
-> **Instant Field-Aligned Meshes**<br/>
-> Wenzel Jakob, Marco Tarini, Daniele Panozzo, Olga Sorkine-Hornung<br/>
-> In *ACM Transactions on Graphics (Proceedings of SIGGRAPH Asia 2015)*<br/>
-> [PDF](http://igl.ethz.ch/projects/instant-meshes/instant-meshes-SA-2015-jakob-et-al.pdf),
-> [Video](https://www.youtube.com/watch?v=U6wtw6W4x3I),
-> [Project page](http://igl.ethz.ch/projects/instant-meshes/)
-
-
-##### In commercial software
-
-Since version 10.2, Modo uses the Instant Meshes algorithm to implement its
-automatic retopology feature. An interview discussing this technique and more
-recent projects is available [here](https://www.foundry.com/trends/design-visualisation/mitsuba-renderer-instant-meshes).
-
-## Screenshot
-
-![Instant Meshes logo](https://github.com/wjakob/instant-meshes/raw/master/resources/screenshot.jpg)
-
-## Pre-compiled binaries
-
-The following binaries (Intel, 64 bit) are automatically generated from the latest GitHub revision.
-
-> [Microsoft Windows](https://instant-meshes.s3.eu-central-1.amazonaws.com/Release/instant-meshes-windows.zip)<br/>
-> [Mac OS X](https://instant-meshes.s3.eu-central-1.amazonaws.com/instant-meshes-macos.zip)<br/>
-> [Linux](https://instant-meshes.s3.eu-central-1.amazonaws.com/instant-meshes-linux.zip)
-
-Please also fetch the following dataset ZIP file and extract it so that the
-``datasets`` folder is in the same directory as ``Instant Meshes``, ``Instant Meshes.app``,
-or ``Instant Meshes.exe``.
-
-> [Datasets](https://instant-meshes.s3.eu-central-1.amazonaws.com/instant-meshes-datasets.zip)
-
-Note: On Linux, Instant Meshes relies on the program ``zenity``, which must be installed.
-
-## Compiling
-
-Compiling from scratch requires CMake and a recent version of XCode on Mac,
-Visual Studio 2015 on Windows, and GCC on Linux. 
-
-On MacOS, compiling should be as simple as
-
-    git clone --recursive https://github.com/wjakob/instant-meshes
-    cd instant-meshes
-    cmake .
-    make -j 4
-
-To build on Linux, please install the prerequisites ``libxrandr-dev``,
-``libxinerama-dev``, ``libxcursor-dev``, and ``libxi-dev`` and then use the
-same sequence of commands shown above for MacOS.
-
-On Windows, open the generated file ``InstantMeshes.sln`` after step 3 and proceed building as usual from within Visual Studio.
-
-## Usage
-
-To get started, launch the binary and select a dataset using the "Open mesh" button on the top left (the application must be located in the same directory as the 'datasets' folder, otherwise the panel will be empty).
-
-The standard workflow is to solve for an orientation field (first blue button) and a position field (second blue button) in sequence, after which the 'Export mesh' button becomes active. Many user interface elements display a descriptive message when hovering the mouse cursor above for a second.
-
-A range of additional information about the input mesh, the computed fields,
-and the output mesh can be visualized using the check boxes accessible via the
-'Advanced' panel.
-
-Clicking the left mouse button and dragging rotates the object; right-dragging
-(or shift+left-dragging) translates, and the mouse wheel zooms. The fields can also be manipulated using brush tools that are accessible by clicking the first icon in each 'Tool' row.

intern/instant-meshes/Readme.Blender.MD

@@ -1 +0,0 @@
-This is a fork of instant-meshes, an alternative to QuadriFlow.

intern/instant-meshes/ext/dset/LICENSE.txt

@@ -1,17 +0,0 @@
-Copyright (c) 2015 Wenzel Jakob <wenzel@inf.ethz.ch>
-
-This software is provided 'as-is', without any express or implied
-warranty.  In no event will the authors be held liable for any damages
-arising from the use of this software.
-
-Permission is granted to anyone to use this software for any purpose,
-including commercial applications, and to alter it and redistribute it
-freely, subject to the following restrictions:
-
-1. The origin of this software must not be misrepresented; you must not
-claim that you wrote the original software. If you use this software
-in a product, an acknowledgment in the product documentation would be
-appreciated but is not required.
-2. Altered source versions must be plainly marked as such, and must not be
-misrepresented as being the original software.
-3. This notice may not be removed or altered from any source distribution.

intern/instant-meshes/ext/dset/README.md

@@ -1,11 +0,0 @@
-# Lock-free parallel disjoint set data structure
-
-This is a small self-contained C++11 implementation of the UNION-FIND data
-structure with path compression and union by rank and a few extras It supports
-concurrent `find()`, `same()` and `unite()` calls as described in the paper
-
-*Wait-free Parallel Algorithms for the Union-Find Problem*
-by Richard J. Anderson and Heather Woll
-
-In addition, this class supports optimistic locking (`try_lock()`/`unlock()`)
-of disjoint sets and a *combined* unite+unlock operation for pairs of sets.

intern/instant-meshes/ext/dset/dset.h

@@ -1,159 +0,0 @@
-#if !defined(__UNIONFIND_H)
-#define __UNIONFIND_H
-
-#include <vector>
-#include <atomic>
-#include <iostream>
-
-/**
- * Lock-free parallel disjoint set data structure (aka UNION-FIND)
- * with path compression and union by rank
- *
- * Supports concurrent find(), same() and unite() calls as described
- * in the paper
- *
- * "Wait-free Parallel Algorithms for the Union-Find Problem"
- * by Richard J. Anderson and Heather Woll
- *
- * In addition, this class supports optimistic locking (try_lock/unlock)
- * of disjoint sets and a combined unite+unlock operation.
- *
- * \author Wenzel Jakob
- */
-class DisjointSets {
-public:
-    DisjointSets(uint32_t size) : mData(size) {
-        for (uint32_t i=0; i<size; ++i)
-            mData[i] = (uint32_t) i;
-    }
-
-    uint32_t find(uint32_t id) const {
-        while (id != parent(id)) {
-            uint64_t value = mData[id];
-            uint32_t new_parent = parent((uint32_t) value);
-            uint64_t new_value =
-                (value & 0xFFFFFFFF00000000ULL) | new_parent;
-            /* Try to update parent (may fail, that's ok) */
-            if (value != new_value)
-                mData[id].compare_exchange_weak(value, new_value);
-            id = new_parent;
-        }
-        return id;
-    }
-
-    bool same(uint32_t id1, uint32_t id2) const {
-        for (;;) {
-            id1 = find(id1);
-            id2 = find(id2);
-            if (id1 == id2)
-                return true;
-            if (parent(id1) == id1)
-                return false;
-        }
-    }
-
-    uint32_t unite(uint32_t id1, uint32_t id2) {
-        for (;;) {
-            id1 = find(id1);
-            id2 = find(id2);
-
-            if (id1 == id2)
-                return id1;
-
-            uint32_t r1 = rank(id1), r2 = rank(id2);
-
-            if (r1 > r2 || (r1 == r2 && id1 < id2)) {
-                std::swap(r1, r2);
-                std::swap(id1, id2);
-            }
-
-            uint64_t oldEntry = ((uint64_t) r1 << 32) | id1;
-            uint64_t newEntry = ((uint64_t) r1 << 32) | id2;
-
-            if (!mData[id1].compare_exchange_strong(oldEntry, newEntry))
-                continue;
-
-            if (r1 == r2) {
-                oldEntry = ((uint64_t) r2 << 32) | id2;
-                newEntry = ((uint64_t) (r2+1) << 32) | id2;
-                /* Try to update the rank (may fail, that's ok) */
-                mData[id2].compare_exchange_weak(oldEntry, newEntry);
-            }
-
-            break;
-        }
-        return id2;
-    }
-
-    /**
-     * Try to lock the a disjoint union identified by one
-     * of its elements (this can occasionally fail when there
-     * are concurrent operations). The parameter 'id' will be
-     * updated to store the current representative ID of the
-     * union
-     */
-    bool try_lock(uint32_t &id) {
-        const uint64_t lock_flag = 1ULL << 63;
-        id = find(id);
-        uint64_t value = mData[id];
-        if ((value & lock_flag) || (uint32_t) value != id)
-            return false;
-        // On IA32/x64, a PAUSE instruction is recommended for CAS busy loops
-        #if defined(__i386__) || defined(__amd64__)
-            __asm__ __volatile__ ("pause\n");
-        #endif
-        return mData[id].compare_exchange_strong(value, value | lock_flag);
-    }
-
-    void unlock(uint32_t id) {
-        const uint64_t lock_flag = 1ULL << 63;
-        mData[id] &= ~lock_flag;
-    }
-
-    /**
-     * Return the representative index of the set that results from merging
-     * locked disjoint sets 'id1' and 'id2'
-     */
-    uint32_t unite_index_locked(uint32_t id1, uint32_t id2) const {
-        uint32_t r1 = rank(id1), r2 = rank(id2);
-        return (r1 > r2 || (r1 == r2 && id1 < id2)) ? id1 : id2;
-    }
-
-    /**
-     * Atomically unite two locked disjoint sets and unlock them. Assumes
-     * that here are no other concurrent unite() involving the same sets
-     */
-    uint32_t unite_unlock(uint32_t id1, uint32_t id2) {
-        uint32_t r1 = rank(id1), r2 = rank(id2);
-
-        if (r1 > r2 || (r1 == r2 && id1 < id2)) {
-            std::swap(r1, r2);
-            std::swap(id1, id2);
-        }
-
-        mData[id1] = ((uint64_t) r1 << 32) | id2;
-        mData[id2] = ((uint64_t) (r2 + ((r1 == r2) ? 1 : 0)) << 32) | id2;
-
-        return id2;
-    }
-
-    uint32_t size() const { return (uint32_t) mData.size(); }
-
-    uint32_t rank(uint32_t id) const {
-        return ((uint32_t) (mData[id] >> 32)) & 0x7FFFFFFFu;
-    }
-
-    uint32_t parent(uint32_t id) const {
-        return (uint32_t) mData[id];
-    }
-
-    friend std::ostream &operator<<(std::ostream &os, const DisjointSets &f) {
-        for (size_t i=0; i<f.mData.size(); ++i)
-            os << i << ": parent=" << f.parent(i) << ", rank=" << f.rank(i) << std::endl;
-        return os;
-    }
-
-    mutable std::vector<std::atomic<uint64_t>> mData;
-};
-
-#endif /* __UNIONFIND_H */