Merge branch 'master' into draw-viewport-data

This is not needed anymore with the new dependency system.

build_files/cmake/platform/platform_unix.cmake

@@ -109,9 +109,14 @@ if(NOT WITH_SYSTEM_FREETYPE)
   find_package_wrapper(Freetype REQUIRED)
   if(EXISTS ${LIBDIR})
     find_package_wrapper(Brotli REQUIRED)
-    list(APPEND FREETYPE_LIBRARIES
-      ${BROTLI_LIBRARIES}
-    )
+
+    # NOTE: This is done on WIN32 & APPLE but fails on some Linux systems.
+    # See: https://devtalk.blender.org/t/22536
+    # So `BROTLI_LIBRARIES` need to be added  `FREETYPE_LIBRARIES`.
+    #
+    # list(APPEND FREETYPE_LIBRARIES
+    #   ${BROTLI_LIBRARIES}
+    # )
   endif()
 endif()
 

intern/cycles/bvh/embree.cpp

@@ -61,6 +61,26 @@ static_assert(Object::MAX_MOTION_STEPS == Geometry::MAX_MOTION_STEPS,
 
 #  define IS_HAIR(x) (x & 1)
 
+/* This gets called by Embree at every valid ray/object intersection.
+ * Things like recording subsurface or shadow hits for later evaluation
+ * as well as filtering for volume objects happen here.
+ * Cycles' own BVH does that directly inside the traversal calls.
+ */
+static void rtc_filter_intersection_func(const RTCFilterFunctionNArguments *args)
+{
+  /* Current implementation in Cycles assumes only single-ray intersection queries. */
+  assert(args->N == 1);
+
+  RTCHit *hit = (RTCHit *)args->hit;
+  CCLIntersectContext *ctx = ((IntersectContext *)args->context)->userRayExt;
+  const KernelGlobalsCPU *kg = ctx->kg;
+  const Ray *cray = ctx->ray;
+
+  if (kernel_embree_is_self_intersection(kg, hit, cray)) {
+    *args->valid = 0;
+  }
+}
+
 /* This gets called by Embree at every valid ray/object intersection.
  * Things like recording subsurface or shadow hits for later evaluation
  * as well as filtering for volume objects happen here.
@@ -75,12 +95,16 @@ static void rtc_filter_occluded_func(const RTCFilterFunctionNArguments *args)
   RTCHit *hit = (RTCHit *)args->hit;
   CCLIntersectContext *ctx = ((IntersectContext *)args->context)->userRayExt;
   const KernelGlobalsCPU *kg = ctx->kg;
+  const Ray *cray = ctx->ray;
 
   switch (ctx->type) {
     case CCLIntersectContext::RAY_SHADOW_ALL: {
       Intersection current_isect;
       kernel_embree_convert_hit(kg, ray, hit, &current_isect);
-
+      if (intersection_skip_self_shadow(cray->self, current_isect.object, current_isect.prim)) {
+        *args->valid = 0;
+        return;
+      }
       /* If no transparent shadows or max number of hits exceeded, all light is blocked. */
       const int flags = intersection_get_shader_flags(kg, current_isect.prim, current_isect.type);
       if (!(flags & (SD_HAS_TRANSPARENT_SHADOW)) || ctx->num_hits >= ctx->max_hits) {
@@ -160,6 +184,10 @@ static void rtc_filter_occluded_func(const RTCFilterFunctionNArguments *args)
           break;
         }
       }
+      if (intersection_skip_self_local(cray->self, current_isect.prim)) {
+        *args->valid = 0;
+        return;
+      }
 
       /* No intersection information requested, just return a hit. */
       if (ctx->max_hits == 0) {
@@ -225,6 +253,11 @@ static void rtc_filter_occluded_func(const RTCFilterFunctionNArguments *args)
       if (ctx->num_hits < ctx->max_hits) {
         Intersection current_isect;
         kernel_embree_convert_hit(kg, ray, hit, &current_isect);
+        if (intersection_skip_self(cray->self, current_isect.object, current_isect.prim)) {
+          *args->valid = 0;
+          return;
+        }
+
         Intersection *isect = &ctx->isect_s[ctx->num_hits];
         ++ctx->num_hits;
         *isect = current_isect;
@@ -236,12 +269,15 @@ static void rtc_filter_occluded_func(const RTCFilterFunctionNArguments *args)
         }
         /* This tells Embree to continue tracing. */
         *args->valid = 0;
-        break;
       }
+      break;
     }
     case CCLIntersectContext::RAY_REGULAR:
     default:
-      /* Nothing to do here. */
+      if (kernel_embree_is_self_intersection(kg, hit, cray)) {
+        *args->valid = 0;
+        return;
+      }
       break;
   }
 }
@@ -257,6 +293,14 @@ static void rtc_filter_func_backface_cull(const RTCFilterFunctionNArguments *arg
     *args->valid = 0;
     return;
   }
+
+  CCLIntersectContext *ctx = ((IntersectContext *)args->context)->userRayExt;
+  const KernelGlobalsCPU *kg = ctx->kg;
+  const Ray *cray = ctx->ray;
+
+  if (kernel_embree_is_self_intersection(kg, hit, cray)) {
+    *args->valid = 0;
+  }
 }
 
 static void rtc_filter_occluded_func_backface_cull(const RTCFilterFunctionNArguments *args)
@@ -505,6 +549,7 @@ void BVHEmbree::add_triangles(const Object *ob, const Mesh *mesh, int i)
 
   rtcSetGeometryUserData(geom_id, (void *)prim_offset);
   rtcSetGeometryOccludedFilterFunction(geom_id, rtc_filter_occluded_func);
+  rtcSetGeometryIntersectFilterFunction(geom_id, rtc_filter_intersection_func);
   rtcSetGeometryMask(geom_id, ob->visibility_for_tracing());
 
   rtcCommitGeometry(geom_id);
@@ -767,6 +812,7 @@ void BVHEmbree::add_curves(const Object *ob, const Hair *hair, int i)
 
   rtcSetGeometryUserData(geom_id, (void *)prim_offset);
   if (hair->curve_shape == CURVE_RIBBON) {
+    rtcSetGeometryIntersectFilterFunction(geom_id, rtc_filter_intersection_func);
     rtcSetGeometryOccludedFilterFunction(geom_id, rtc_filter_occluded_func);
   }
   else {

intern/cycles/device/optix/device_impl.cpp

@@ -226,7 +226,7 @@ bool OptiXDevice::load_kernels(const uint kernel_features)
   pipeline_options.usesMotionBlur = false;
   pipeline_options.traversableGraphFlags =
       OPTIX_TRAVERSABLE_GRAPH_FLAG_ALLOW_SINGLE_LEVEL_INSTANCING;
-  pipeline_options.numPayloadValues = 6;
+  pipeline_options.numPayloadValues = 8;
   pipeline_options.numAttributeValues = 2; /* u, v */
   pipeline_options.exceptionFlags = OPTIX_EXCEPTION_FLAG_NONE;
   pipeline_options.pipelineLaunchParamsVariableName = "__params"; /* See globals.h */

intern/cycles/kernel/bvh/bvh.h

@@ -173,15 +173,16 @@ ccl_device_intersect bool scene_intersect(KernelGlobals kg,
   uint p3 = 0;
   uint p4 = visibility;
   uint p5 = PRIMITIVE_NONE;
+  uint p6 = ((uint64_t)ray) & 0xFFFFFFFF;
+  uint p7 = (((uint64_t)ray) >> 32) & 0xFFFFFFFF;
 
   uint ray_mask = visibility & 0xFF;
-  uint ray_flags = OPTIX_RAY_FLAG_NONE;
+  uint ray_flags = OPTIX_RAY_FLAG_ENFORCE_ANYHIT;
   if (0 == ray_mask && (visibility & ~0xFF) != 0) {
     ray_mask = 0xFF;
-    ray_flags = OPTIX_RAY_FLAG_ENFORCE_ANYHIT;
   }
   else if (visibility & PATH_RAY_SHADOW_OPAQUE) {
-    ray_flags = OPTIX_RAY_FLAG_TERMINATE_ON_FIRST_HIT;
+    ray_flags |= OPTIX_RAY_FLAG_TERMINATE_ON_FIRST_HIT;
   }
 
   optixTrace(scene_intersect_valid(ray) ? kernel_data.bvh.scene : 0,
@@ -200,7 +201,9 @@ ccl_device_intersect bool scene_intersect(KernelGlobals kg,
              p2,
              p3,
              p4,
-             p5);
+             p5,
+             p6,
+             p7);
 
   isect->t = __uint_as_float(p0);
   isect->u = __uint_as_float(p1);
@@ -242,6 +245,7 @@ ccl_device_intersect bool scene_intersect(KernelGlobals kg,
   }
 
   MetalRTIntersectionPayload payload;
+  payload.self = ray->self;
   payload.u = 0.0f;
   payload.v = 0.0f;
   payload.visibility = visibility;
@@ -309,6 +313,7 @@ ccl_device_intersect bool scene_intersect(KernelGlobals kg,
     CCLIntersectContext ctx(kg, CCLIntersectContext::RAY_REGULAR);
     IntersectContext rtc_ctx(&ctx);
     RTCRayHit ray_hit;
+    ctx.ray = ray;
     kernel_embree_setup_rayhit(*ray, ray_hit, visibility);
     rtcIntersect1(kernel_data.bvh.scene, &rtc_ctx.context, &ray_hit);
     if (ray_hit.hit.geomID != RTC_INVALID_GEOMETRY_ID &&
@@ -356,6 +361,9 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
   uint p2 = pointer_pack_to_uint_0(local_isect);
   uint p3 = pointer_pack_to_uint_1(local_isect);
   uint p4 = local_object;
+  uint p6 = ((uint64_t)ray) & 0xFFFFFFFF;
+  uint p7 = (((uint64_t)ray) >> 32) & 0xFFFFFFFF;
+
   /* Is set to zero on miss or if ray is aborted, so can be used as return value. */
   uint p5 = max_hits;
 
@@ -379,7 +387,9 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
              p2,
              p3,
              p4,
-             p5);
+             p5,
+             p6,
+             p7);
 
   return p5;
 #  elif defined(__METALRT__)
@@ -417,6 +427,7 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
   }
 
   MetalRTIntersectionLocalPayload payload;
+  payload.self = ray->self;
   payload.local_object = local_object;
   payload.max_hits = max_hits;
   payload.local_isect.num_hits = 0;
@@ -460,6 +471,7 @@ ccl_device_intersect bool scene_intersect_local(KernelGlobals kg,
         kg, has_bvh ? CCLIntersectContext::RAY_SSS : CCLIntersectContext::RAY_LOCAL);
     ctx.lcg_state = lcg_state;
     ctx.max_hits = max_hits;
+    ctx.ray = ray;
     ctx.local_isect = local_isect;
     if (local_isect) {
       local_isect->num_hits = 0;
@@ -532,6 +544,8 @@ ccl_device_intersect bool scene_intersect_shadow_all(KernelGlobals kg,
   uint p3 = max_hits;
   uint p4 = visibility;
   uint p5 = false;
+  uint p6 = ((uint64_t)ray) & 0xFFFFFFFF;
+  uint p7 = (((uint64_t)ray) >> 32) & 0xFFFFFFFF;
 
   uint ray_mask = visibility & 0xFF;
   if (0 == ray_mask && (visibility & ~0xFF) != 0) {
@@ -555,7 +569,9 @@ ccl_device_intersect bool scene_intersect_shadow_all(KernelGlobals kg,
              p2,
              p3,
              p4,
-             p5);
+             p5,
+             p6,
+             p7);
 
   *num_recorded_hits = uint16_unpack_from_uint_0(p2);
   *throughput = __uint_as_float(p1);
@@ -588,6 +604,7 @@ ccl_device_intersect bool scene_intersect_shadow_all(KernelGlobals kg,
   }
 
   MetalRTIntersectionShadowPayload payload;
+  payload.self = ray->self;
   payload.visibility = visibility;
   payload.max_hits = max_hits;
   payload.num_hits = 0;
@@ -634,6 +651,7 @@ ccl_device_intersect bool scene_intersect_shadow_all(KernelGlobals kg,
     Intersection *isect_array = (Intersection *)state->shadow_isect;
     ctx.isect_s = isect_array;
     ctx.max_hits = max_hits;
+    ctx.ray = ray;
     IntersectContext rtc_ctx(&ctx);
     RTCRay rtc_ray;
     kernel_embree_setup_ray(*ray, rtc_ray, visibility);
@@ -685,6 +703,8 @@ ccl_device_intersect bool scene_intersect_volume(KernelGlobals kg,
   uint p3 = 0;
   uint p4 = visibility;
   uint p5 = PRIMITIVE_NONE;
+  uint p6 = ((uint64_t)ray) & 0xFFFFFFFF;
+  uint p7 = (((uint64_t)ray) >> 32) & 0xFFFFFFFF;
 
   uint ray_mask = visibility & 0xFF;
   if (0 == ray_mask && (visibility & ~0xFF) != 0) {
@@ -708,7 +728,9 @@ ccl_device_intersect bool scene_intersect_volume(KernelGlobals kg,
              p2,
              p3,
              p4,
-             p5);
+             p5,
+             p6,
+             p7);
 
   isect->t = __uint_as_float(p0);
   isect->u = __uint_as_float(p1);
@@ -744,6 +766,7 @@ ccl_device_intersect bool scene_intersect_volume(KernelGlobals kg,
   }
 
   MetalRTIntersectionPayload payload;
+  payload.self = ray->self;
   payload.visibility = visibility;
 
   typename metalrt_intersector_type::result_type intersection;
@@ -820,6 +843,7 @@ ccl_device_intersect uint scene_intersect_volume_all(KernelGlobals kg,
     ctx.isect_s = isect;
     ctx.max_hits = max_hits;
     ctx.num_hits = 0;
+    ctx.ray = ray;
     IntersectContext rtc_ctx(&ctx);
     RTCRay rtc_ray;
     kernel_embree_setup_ray(*ray, rtc_ray, visibility);

intern/cycles/kernel/bvh/embree.h

@@ -22,6 +22,8 @@
 #include "kernel/device/cpu/compat.h"
 #include "kernel/device/cpu/globals.h"
 
+#include "kernel/bvh/util.h"
+
 #include "util/vector.h"
 
 CCL_NAMESPACE_BEGIN
@@ -38,6 +40,9 @@ struct CCLIntersectContext {
   KernelGlobals kg;
   RayType type;
 
+  /* For avoiding self intersections */
+  const Ray *ray;
+
   /* for shadow rays */
   Intersection *isect_s;
   uint max_hits;
@@ -56,6 +61,7 @@ struct CCLIntersectContext {
   {
     kg = kg_;
     type = type_;
+    ray = NULL;
     max_hits = 1;
     num_hits = 0;
     num_recorded_hits = 0;
@@ -102,7 +108,34 @@ ccl_device_inline void kernel_embree_setup_rayhit(const Ray &ray,
 {
   kernel_embree_setup_ray(ray, rayhit.ray, visibility);
   rayhit.hit.geomID = RTC_INVALID_GEOMETRY_ID;
-  rayhit.hit.primID = RTC_INVALID_GEOMETRY_ID;
+  rayhit.hit.instID[0] = RTC_INVALID_GEOMETRY_ID;
+}
+
+ccl_device_inline bool kernel_embree_is_self_intersection(const KernelGlobals kg,
+                                                          const RTCHit *hit,
+                                                          const Ray *ray)
+{
+  bool status = false;
+  if (hit->instID[0] != RTC_INVALID_GEOMETRY_ID) {
+    const int oID = hit->instID[0] / 2;
+    if ((ray->self.object == oID) || (ray->self.light_object == oID)) {
+      RTCScene inst_scene = (RTCScene)rtcGetGeometryUserData(
+          rtcGetGeometry(kernel_data.bvh.scene, hit->instID[0]));
+      const int pID = hit->primID +
+                      (intptr_t)rtcGetGeometryUserData(rtcGetGeometry(inst_scene, hit->geomID));
+      status = intersection_skip_self_shadow(ray->self, oID, pID);
+    }
+  }
+  else {
+    const int oID = hit->geomID / 2;
+    if ((ray->self.object == oID) || (ray->self.light_object == oID)) {
+      const int pID = hit->primID + (intptr_t)rtcGetGeometryUserData(
+                                        rtcGetGeometry(kernel_data.bvh.scene, hit->geomID));
+      status = intersection_skip_self_shadow(ray->self, oID, pID);
+    }
+  }
+
+  return status;
 }
 
 ccl_device_inline void kernel_embree_convert_hit(KernelGlobals kg,

intern/cycles/kernel/bvh/local.h

@@ -157,7 +157,11 @@ ccl_device_inline
                 }
               }
 
+              /* Skip self intersection. */
               const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+              if (intersection_skip_self_local(ray->self, prim)) {
+                continue;
+              }
 
               if (triangle_intersect_local(kg,
                                            local_isect,
@@ -188,7 +192,11 @@ ccl_device_inline
                 }
               }
 
+              /* Skip self intersection. */
               const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+              if (intersection_skip_self_local(ray->self, prim)) {
+                continue;
+              }
 
               if (motion_triangle_intersect_local(kg,
                                                   local_isect,

intern/cycles/kernel/bvh/metal.h

@@ -15,6 +15,7 @@
  */
 
 struct MetalRTIntersectionPayload {
+  RaySelfPrimitives self;
   uint visibility;
   float u, v;
   int prim;
@@ -25,6 +26,7 @@ struct MetalRTIntersectionPayload {
 };
 
 struct MetalRTIntersectionLocalPayload {
+  RaySelfPrimitives self;
   uint local_object;
   uint lcg_state;
   short max_hits;
@@ -34,6 +36,7 @@ struct MetalRTIntersectionLocalPayload {
 };
 
 struct MetalRTIntersectionShadowPayload {
+  RaySelfPrimitives self;
   uint visibility;
 #if defined(__METALRT_MOTION__)
   float time;

intern/cycles/kernel/bvh/shadow_all.h

@@ -160,6 +160,9 @@ ccl_device_inline
                                         kernel_tex_fetch(__prim_object, prim_addr) :
                                         object;
             const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+            if (intersection_skip_self_shadow(ray->self, prim_object, prim)) {
+              continue;
+            }
 
             switch (type & PRIMITIVE_ALL) {
               case PRIMITIVE_TRIANGLE: {

intern/cycles/kernel/bvh/traversal.h

@@ -133,35 +133,29 @@ ccl_device_noinline bool BVH_FUNCTION_FULL_NAME(BVH)(KernelGlobals kg,
           --stack_ptr;
 
           /* primitive intersection */
-          switch (type & PRIMITIVE_ALL) {
-            case PRIMITIVE_TRIANGLE: {
-              for (; prim_addr < prim_addr2; prim_addr++) {
-                kernel_assert(kernel_tex_fetch(__prim_type, prim_addr) == type);
+          for (; prim_addr < prim_addr2; prim_addr++) {
+            kernel_assert(kernel_tex_fetch(__prim_type, prim_addr) == type);
 
-                const int prim_object = (object == OBJECT_NONE) ?
-                                            kernel_tex_fetch(__prim_object, prim_addr) :
-                                            object;
-                const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+            const int prim_object = (object == OBJECT_NONE) ?
+                                        kernel_tex_fetch(__prim_object, prim_addr) :
+                                        object;
+            const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+            if (intersection_skip_self_shadow(ray->self, prim_object, prim)) {
+              continue;
+            }
 
+            switch (type & PRIMITIVE_ALL) {
+              case PRIMITIVE_TRIANGLE: {
                 if (triangle_intersect(
                         kg, isect, P, dir, isect->t, visibility, prim_object, prim, prim_addr)) {
                   /* shadow ray early termination */
                   if (visibility & PATH_RAY_SHADOW_OPAQUE)
                     return true;
                 }
+                break;
               }
-              break;
-            }
 #if BVH_FEATURE(BVH_MOTION)
-            case PRIMITIVE_MOTION_TRIANGLE: {
-              for (; prim_addr < prim_addr2; prim_addr++) {
-                kernel_assert(kernel_tex_fetch(__prim_type, prim_addr) == type);
-
-                const int prim_object = (object == OBJECT_NONE) ?
-                                            kernel_tex_fetch(__prim_object, prim_addr) :
-                                            object;
-                const int prim = kernel_tex_fetch(__prim_index, prim_addr);
-
+              case PRIMITIVE_MOTION_TRIANGLE: {
                 if (motion_triangle_intersect(kg,
                                               isect,
                                               P,
@@ -176,28 +170,21 @@ ccl_device_noinline bool BVH_FUNCTION_FULL_NAME(BVH)(KernelGlobals kg,
                   if (visibility & PATH_RAY_SHADOW_OPAQUE)
                     return true;
                 }
+                break;
               }
-              break;
-            }
 #endif /* BVH_FEATURE(BVH_MOTION) */
 #if BVH_FEATURE(BVH_HAIR)
-            case PRIMITIVE_CURVE_THICK:
-            case PRIMITIVE_MOTION_CURVE_THICK:
-            case PRIMITIVE_CURVE_RIBBON:
-            case PRIMITIVE_MOTION_CURVE_RIBBON: {
-              for (; prim_addr < prim_addr2; prim_addr++) {
+              case PRIMITIVE_CURVE_THICK:
+              case PRIMITIVE_MOTION_CURVE_THICK:
+              case PRIMITIVE_CURVE_RIBBON:
+              case PRIMITIVE_MOTION_CURVE_RIBBON: {
                 if ((type & PRIMITIVE_MOTION) && kernel_data.bvh.use_bvh_steps) {
                   const float2 prim_time = kernel_tex_fetch(__prim_time, prim_addr);
                   if (ray->time < prim_time.x || ray->time > prim_time.y) {
-                    continue;
+                    break;
                   }
                 }
 
-                const int prim_object = (object == OBJECT_NONE) ?
-                                            kernel_tex_fetch(__prim_object, prim_addr) :
-                                            object;
-                const int prim = kernel_tex_fetch(__prim_index, prim_addr);
-
                 const int curve_type = kernel_tex_fetch(__prim_type, prim_addr);
                 const bool hit = curve_intersect(
                     kg, isect, P, dir, isect->t, prim_object, prim, ray->time, curve_type);
@@ -206,26 +193,19 @@ ccl_device_noinline bool BVH_FUNCTION_FULL_NAME(BVH)(KernelGlobals kg,
                   if (visibility & PATH_RAY_SHADOW_OPAQUE)
                     return true;
                 }
+                break;
               }
-              break;
-            }
 #endif /* BVH_FEATURE(BVH_HAIR) */
 #if BVH_FEATURE(BVH_POINTCLOUD)
-            case PRIMITIVE_POINT:
-            case PRIMITIVE_MOTION_POINT: {
-              for (; prim_addr < prim_addr2; prim_addr++) {
+              case PRIMITIVE_POINT:
+              case PRIMITIVE_MOTION_POINT: {
                 if ((type & PRIMITIVE_MOTION) && kernel_data.bvh.use_bvh_steps) {
                   const float2 prim_time = kernel_tex_fetch(__prim_time, prim_addr);
                   if (ray->time < prim_time.x || ray->time > prim_time.y) {
-                    continue;
+                    break;
                   }
                 }
 
-                const int prim_object = (object == OBJECT_NONE) ?
-                                            kernel_tex_fetch(__prim_object, prim_addr) :
-                                            object;
-                const int prim = kernel_tex_fetch(__prim_index, prim_addr);
-
                 const int point_type = kernel_tex_fetch(__prim_type, prim_addr);
                 const bool hit = point_intersect(
                     kg, isect, P, dir, isect->t, prim_object, prim, ray->time, point_type);
@@ -234,10 +214,10 @@ ccl_device_noinline bool BVH_FUNCTION_FULL_NAME(BVH)(KernelGlobals kg,
                   if (visibility & PATH_RAY_SHADOW_OPAQUE)
                     return true;
                 }
+                break;
               }
-              break;
-            }
 #endif /* BVH_FEATURE(BVH_POINTCLOUD) */
+            }
           }
         }
         else {

intern/cycles/kernel/bvh/util.h

@@ -21,54 +21,22 @@ CCL_NAMESPACE_BEGIN
 /* Ray offset to avoid self intersection.
  *
  * This function should be used to compute a modified ray start position for
- * rays leaving from a surface. */
-
+ * rays leaving from a surface. This is from "A Fast and Robust Method for Avoiding
+ * Self-Intersection" see https://research.nvidia.com/publication/2019-03_A-Fast-and
+ */
 ccl_device_inline float3 ray_offset(float3 P, float3 Ng)
 {
-#ifdef __INTERSECTION_REFINE__
-  const float epsilon_f = 1e-5f;
-  /* ideally this should match epsilon_f, but instancing and motion blur
-   * precision makes it problematic */
-  const float epsilon_test = 1.0f;
-  const int epsilon_i = 32;
+  const float int_scale = 256.0f;
+  int3 of_i = make_int3((int)(int_scale * Ng.x), (int)(int_scale * Ng.y), (int)(int_scale * Ng.z));
 
-  float3 res;
-
-  /* x component */
-  if (fabsf(P.x) < epsilon_test) {
-    res.x = P.x + Ng.x * epsilon_f;
-  }
-  else {
-    uint ix = __float_as_uint(P.x);
-    ix += ((ix ^ __float_as_uint(Ng.x)) >> 31) ? -epsilon_i : epsilon_i;
-    res.x = __uint_as_float(ix);
-  }
-
-  /* y component */
-  if (fabsf(P.y) < epsilon_test) {
-    res.y = P.y + Ng.y * epsilon_f;
-  }
-  else {
-    uint iy = __float_as_uint(P.y);
-    iy += ((iy ^ __float_as_uint(Ng.y)) >> 31) ? -epsilon_i : epsilon_i;
-    res.y = __uint_as_float(iy);
-  }
-
-  /* z component */
-  if (fabsf(P.z) < epsilon_test) {
-    res.z = P.z + Ng.z * epsilon_f;
-  }
-  else {
-    uint iz = __float_as_uint(P.z);
-    iz += ((iz ^ __float_as_uint(Ng.z)) >> 31) ? -epsilon_i : epsilon_i;
-    res.z = __uint_as_float(iz);
-  }
-
-  return res;
-#else
-  const float epsilon_f = 1e-4f;
-  return P + epsilon_f * Ng;
-#endif
+  float3 p_i = make_float3(__int_as_float(__float_as_int(P.x) + ((P.x < 0) ? -of_i.x : of_i.x)),
+                           __int_as_float(__float_as_int(P.y) + ((P.y < 0) ? -of_i.y : of_i.y)),
+                           __int_as_float(__float_as_int(P.z) + ((P.z < 0) ? -of_i.z : of_i.z)));
+  const float origin = 1.0f / 32.0f;
+  const float float_scale = 1.0f / 65536.0f;
+  return make_float3(fabsf(P.x) < origin ? P.x + float_scale * Ng.x : p_i.x,
+                     fabsf(P.y) < origin ? P.y + float_scale * Ng.y : p_i.y,
+                     fabsf(P.z) < origin ? P.z + float_scale * Ng.z : p_i.z);
 }
 
 #if defined(__KERNEL_CPU__)
@@ -227,4 +195,25 @@ ccl_device_inline float intersection_curve_shadow_transparency(KernelGlobals kg,
   return (1.0f - u) * f0 + u * f1;
 }
 
+ccl_device_inline bool intersection_skip_self(ccl_private const RaySelfPrimitives &self,
+                                              const int object,
+                                              const int prim)
+{
+  return (self.prim == prim) && (self.object == object);
+}
+
+ccl_device_inline bool intersection_skip_self_shadow(ccl_private const RaySelfPrimitives &self,
+                                                     const int object,
+                                                     const int prim)
+{
+  return ((self.prim == prim) && (self.object == object)) ||
+         ((self.light_prim == prim) && (self.light_object == object));
+}
+
+ccl_device_inline bool intersection_skip_self_local(ccl_private const RaySelfPrimitives &self,
+                                                    const int prim)
+{
+  return (self.prim == prim);
+}
+
 CCL_NAMESPACE_END

intern/cycles/kernel/bvh/volume.h

@@ -144,6 +144,9 @@ ccl_device_inline
                                             kernel_tex_fetch(__prim_object, prim_addr) :
                                             object;
                 const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+                if (intersection_skip_self(ray->self, prim_object, prim)) {
+                  continue;
+                }
 
                 int object_flag = kernel_tex_fetch(__object_flag, prim_object);
                 if ((object_flag & SD_OBJECT_HAS_VOLUME) == 0) {
@@ -164,6 +167,9 @@ ccl_device_inline
                                             kernel_tex_fetch(__prim_object, prim_addr) :
                                             object;
                 const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+                if (intersection_skip_self(ray->self, prim_object, prim)) {
+                  continue;
+                }
                 int object_flag = kernel_tex_fetch(__object_flag, prim_object);
                 if ((object_flag & SD_OBJECT_HAS_VOLUME) == 0) {
                   continue;

intern/cycles/kernel/bvh/volume_all.h

@@ -147,6 +147,9 @@ ccl_device_inline
                                             kernel_tex_fetch(__prim_object, prim_addr) :
                                             object;
                 const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+                if (intersection_skip_self(ray->self, prim_object, prim)) {
+                  continue;
+                }
                 int object_flag = kernel_tex_fetch(__object_flag, prim_object);
                 if ((object_flag & SD_OBJECT_HAS_VOLUME) == 0) {
                   continue;
@@ -188,6 +191,9 @@ ccl_device_inline
                                             kernel_tex_fetch(__prim_object, prim_addr) :
                                             object;
                 const int prim = kernel_tex_fetch(__prim_index, prim_addr);
+                if (intersection_skip_self(ray->self, prim_object, prim)) {
+                  continue;
+                }
                 int object_flag = kernel_tex_fetch(__object_flag, prim_object);
                 if ((object_flag & SD_OBJECT_HAS_VOLUME) == 0) {
                   continue;

intern/cycles/kernel/device/metal/kernel.metal

@@ -40,6 +40,27 @@ struct TriangleIntersectionResult
 
 enum { METALRT_HIT_TRIANGLE, METALRT_HIT_BOUNDING_BOX };
 
+ccl_device_inline bool intersection_skip_self(ray_data const RaySelfPrimitives& self,
+                                              const int object,
+                                              const int prim)
+{
+  return (self.prim == prim) && (self.object == object);
+}
+
+ccl_device_inline bool intersection_skip_self_shadow(ray_data const RaySelfPrimitives& self,
+                                                     const int object,
+                                                     const int prim)
+{
+  return ((self.prim == prim) && (self.object == object)) ||
+         ((self.light_prim == prim) && (self.light_object == object));
+}
+
+ccl_device_inline bool intersection_skip_self_local(ray_data const RaySelfPrimitives& self,
+                                                    const int prim)
+{
+  return (self.prim == prim);
+}
+
 template<typename TReturn, uint intersection_type>
 TReturn metalrt_local_hit(constant KernelParamsMetal &launch_params_metal,
                           ray_data MetalKernelContext::MetalRTIntersectionLocalPayload &payload,
@@ -53,8 +74,8 @@ TReturn metalrt_local_hit(constant KernelParamsMetal &launch_params_metal,
 #ifdef __BVH_LOCAL__
   uint prim = primitive_id + kernel_tex_fetch(__object_prim_offset, object);
 
-  if (object != payload.local_object) {
-    /* Only intersect with matching object */
+  if ((object != payload.local_object) || intersection_skip_self_local(payload.self, prim)) {
+    /* Only intersect with matching object and skip self-intersecton. */
     result.accept = false;
     result.continue_search = true;
     return result;
@@ -166,6 +187,11 @@ bool metalrt_shadow_all_hit(constant KernelParamsMetal &launch_params_metal,
   }
 #  endif
 
+  if (intersection_skip_self_shadow(payload.self, object, prim)) {
+    /* continue search */
+    return true;
+  }
+
   float u = 0.0f, v = 0.0f;
   int type = 0;
   if (intersection_type == METALRT_HIT_TRIANGLE) {
@@ -322,21 +348,35 @@ inline TReturnType metalrt_visibility_test(constant KernelParamsMetal &launch_pa
   }
 #  endif
 
-#  ifdef __VISIBILITY_FLAG__
   uint visibility = payload.visibility;
+#  ifdef __VISIBILITY_FLAG__
   if ((kernel_tex_fetch(__objects, object).visibility & visibility) == 0) {
     result.accept = false;
     result.continue_search = true;
     return result;
   }
+#  endif
 
   /* Shadow ray early termination. */
   if (visibility & PATH_RAY_SHADOW_OPAQUE) {
-    result.accept = true;
-    result.continue_search = false;
-    return result;
+    if (intersection_skip_self_shadow(payload.self, object, prim)) {
+      result.accept = false;
+      result.continue_search = true;
+      return result;
+    }
+    else {
+      result.accept = true;
+      result.continue_search = false;
+      return result;
+    }
+  }
+  else {
+    if (intersection_skip_self(payload.self, object, prim)) {
+      result.accept = false;
+      result.continue_search = true;
+      return result;
+    }
   }
-#  endif
 
   result.accept = true;
   result.continue_search = true;

intern/cycles/kernel/device/optix/kernel.cu

@@ -45,6 +45,11 @@ template<typename T> ccl_device_forceinline T *get_payload_ptr_2()
   return pointer_unpack_from_uint<T>(optixGetPayload_2(), optixGetPayload_3());
 }
 
+template<typename T> ccl_device_forceinline T *get_payload_ptr_6()
+{
+  return (T *)(((uint64_t)optixGetPayload_7() << 32) | optixGetPayload_6());
+}
+
 ccl_device_forceinline int get_object_id()
 {
 #ifdef __OBJECT_MOTION__
@@ -111,6 +116,12 @@ extern "C" __global__ void __anyhit__kernel_optix_local_hit()
     return optixIgnoreIntersection();
   }
 
+  const int prim = optixGetPrimitiveIndex();
+  ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
+  if (intersection_skip_self_local(ray->self, prim)) {
+    return optixIgnoreIntersection();
+  }
+
   const uint max_hits = optixGetPayload_5();
   if (max_hits == 0) {
     /* Special case for when no hit information is requested, just report that something was hit */
@@ -149,8 +160,6 @@ extern "C" __global__ void __anyhit__kernel_optix_local_hit()
     local_isect->num_hits = 1;
   }
 
-  const int prim = optixGetPrimitiveIndex();
-
   Intersection *isect = &local_isect->hits[hit];
   isect->t = optixGetRayTmax();
   isect->prim = prim;
@@ -185,6 +194,11 @@ extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit()
   }
 #  endif
 
+  ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
+  if (intersection_skip_self_shadow(ray->self, object, prim)) {
+    return optixIgnoreIntersection();
+  }
+
   float u = 0.0f, v = 0.0f;
   int type = 0;
   if (optixIsTriangleHit()) {
@@ -314,6 +328,12 @@ extern "C" __global__ void __anyhit__kernel_optix_volume_test()
   if ((kernel_tex_fetch(__object_flag, object) & SD_OBJECT_HAS_VOLUME) == 0) {
     return optixIgnoreIntersection();
   }
+
+  const int prim = optixGetPrimitiveIndex();
+  ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
+  if (intersection_skip_self(ray->self, object, prim)) {
+    return optixIgnoreIntersection();
+  }
 }
 
 extern "C" __global__ void __anyhit__kernel_optix_visibility_test()
@@ -330,18 +350,31 @@ extern "C" __global__ void __anyhit__kernel_optix_visibility_test()
 #  endif
 #endif
 
-#ifdef __VISIBILITY_FLAG__
   const uint object = get_object_id();
   const uint visibility = optixGetPayload_4();
+#ifdef __VISIBILITY_FLAG__
   if ((kernel_tex_fetch(__objects, object).visibility & visibility) == 0) {
     return optixIgnoreIntersection();
   }
-
-  /* Shadow ray early termination. */
-  if (visibility & PATH_RAY_SHADOW_OPAQUE) {
-    return optixTerminateRay();
-  }
 #endif
+
+  const int prim = optixGetPrimitiveIndex();
+  ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
+
+  if (visibility & PATH_RAY_SHADOW_OPAQUE) {
+    if (intersection_skip_self_shadow(ray->self, object, prim)) {
+      return optixIgnoreIntersection();
+    }
+    else {
+      /* Shadow ray early termination. */
+      return optixTerminateRay();
+    }
+  }
+  else {
+    if (intersection_skip_self(ray->self, object, prim)) {
+      return optixIgnoreIntersection();
+    }
+  }
 }
 
 extern "C" __global__ void __closesthit__kernel_optix_hit()

intern/cycles/kernel/geom/motion_triangle_intersect.h

@@ -29,46 +29,19 @@
 
 CCL_NAMESPACE_BEGIN
 
-/* Refine triangle intersection to more precise hit point. For rays that travel
- * far the precision is often not so good, this reintersects the primitive from
- * a closer distance.
+/**
+ * Use the barycentric coordinates to get the intersection location
  */
-
-ccl_device_inline float3 motion_triangle_refine(KernelGlobals kg,
-                                                ccl_private ShaderData *sd,
-                                                float3 P,
-                                                float3 D,
-                                                float t,
-                                                const int isect_object,
-                                                const int isect_prim,
-                                                float3 verts[3])
+ccl_device_inline float3 motion_triangle_point_from_uv(KernelGlobals kg,
+                                                       ccl_private ShaderData *sd,
+                                                       const int isect_object,
+                                                       const int isect_prim,
+                                                       const float u,
+                                                       const float v,
+                                                       float3 verts[3])
 {
-#ifdef __INTERSECTION_REFINE__
-  if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
-    if (UNLIKELY(t == 0.0f)) {
-      return P;
-    }
-    const Transform tfm = object_get_inverse_transform(kg, sd);
-
-    P = transform_point(&tfm, P);
-    D = transform_direction(&tfm, D * t);
-    D = normalize_len(D, &t);
-  }
-
-  P = P + D * t;
-
-  /* Compute refined intersection distance. */
-  const float3 e1 = verts[0] - verts[2];
-  const float3 e2 = verts[1] - verts[2];
-  const float3 s1 = cross(D, e2);
-
-  const float invdivisor = 1.0f / dot(s1, e1);
-  const float3 d = P - verts[2];
-  const float3 s2 = cross(d, e1);
-  float rt = dot(e2, s2) * invdivisor;
-
-  /* Compute refined position. */
-  P = P + D * rt;
+  float w = 1.0f - u - v;
+  float3 P = u * verts[0] + v * verts[1] + w * verts[2];
 
   if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
     const Transform tfm = object_get_transform(kg, sd);
@@ -76,71 +49,8 @@ ccl_device_inline float3 motion_triangle_refine(KernelGlobals kg,
   }
 
   return P;
-#else
-  return P + D * t;
-#endif
 }
 
-/* Same as above, except that t is assumed to be in object space
- * for instancing.
- */
-
-#ifdef __BVH_LOCAL__
-#  if defined(__KERNEL_CUDA__) && (defined(i386) || defined(_M_IX86))
-ccl_device_noinline
-#  else
-ccl_device_inline
-#  endif
-    float3
-    motion_triangle_refine_local(KernelGlobals kg,
-                                 ccl_private ShaderData *sd,
-                                 float3 P,
-                                 float3 D,
-                                 float t,
-                                 const int isect_object,
-                                 const int isect_prim,
-                                 float3 verts[3])
-{
-#  if defined(__KERNEL_GPU_RAYTRACING__)
-  /* t is always in world space with OptiX and MetalRT. */
-  return motion_triangle_refine(kg, sd, P, D, t, isect_object, isect_prim, verts);
-#  else
-#    ifdef __INTERSECTION_REFINE__
-  if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
-    const Transform tfm = object_get_inverse_transform(kg, sd);
-
-    P = transform_point(&tfm, P);
-    D = transform_direction(&tfm, D);
-    D = normalize(D);
-  }
-
-  P = P + D * t;
-
-  /* compute refined intersection distance */
-  const float3 e1 = verts[0] - verts[2];
-  const float3 e2 = verts[1] - verts[2];
-  const float3 s1 = cross(D, e2);
-
-  const float invdivisor = 1.0f / dot(s1, e1);
-  const float3 d = P - verts[2];
-  const float3 s2 = cross(d, e1);
-  float rt = dot(e2, s2) * invdivisor;
-
-  P = P + D * rt;
-
-  if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
-    const Transform tfm = object_get_transform(kg, sd);
-    P = transform_point(&tfm, P);
-  }
-
-  return P;
-#    else  /* __INTERSECTION_REFINE__ */
-  return P + D * t;
-#    endif /* __INTERSECTION_REFINE__ */
-#  endif
-}
-#endif /* __BVH_LOCAL__ */
-
 /* Ray intersection. We simply compute the vertex positions at the given ray
  * time and do a ray intersection with the resulting triangle.
  */

intern/cycles/kernel/geom/motion_triangle_shader.h

@@ -68,15 +68,7 @@ ccl_device_noinline void motion_triangle_shader_setup(KernelGlobals kg,
   verts[1] = (1.0f - t) * verts[1] + t * next_verts[1];
   verts[2] = (1.0f - t) * verts[2] + t * next_verts[2];
   /* Compute refined position. */
-#ifdef __BVH_LOCAL__
-  if (is_local) {
-    sd->P = motion_triangle_refine_local(kg, sd, P, D, ray_t, isect_object, isect_prim, verts);
-  }
-  else
-#endif /* __BVH_LOCAL__*/
-  {
-    sd->P = motion_triangle_refine(kg, sd, P, D, ray_t, isect_object, isect_prim, verts);
-  }
+  sd->P = motion_triangle_point_from_uv(kg, sd, isect_object, isect_prim, sd->u, sd->v, verts);
   /* Compute face normal. */
   float3 Ng;
   if (sd->object_flag & SD_OBJECT_NEGATIVE_SCALE_APPLIED) {

intern/cycles/kernel/geom/shader_data.h

@@ -89,7 +89,7 @@ ccl_device_inline void shader_setup_from_ray(KernelGlobals kg,
       sd->shader = kernel_tex_fetch(__tri_shader, sd->prim);
 
       /* vectors */
-      sd->P = triangle_refine(kg, sd, ray->P, ray->D, isect->t, isect->object, isect->prim);
+      sd->P = triangle_point_from_uv(kg, sd, isect->object, isect->prim, isect->u, isect->v);
       sd->Ng = Ng;
       sd->N = Ng;
 

intern/cycles/kernel/geom/triangle_intersect.h

@@ -142,58 +142,23 @@ ccl_device_inline bool triangle_intersect_local(KernelGlobals kg,
 }
 #endif /* __BVH_LOCAL__ */
 
-/* Refine triangle intersection to more precise hit point. For rays that travel
- * far the precision is often not so good, this reintersects the primitive from
- * a closer distance. */
-
-/* Reintersections uses the paper:
- *
- * Tomas Moeller
- * Fast, minimum storage ray/triangle intersection
- * http://www.cs.virginia.edu/~gfx/Courses/2003/ImageSynthesis/papers/Acceleration/Fast%20MinimumStorage%20RayTriangle%20Intersection.pdf
+/**
+ * Use the barycentric coordinates to get the intersection location
  */
-
-ccl_device_inline float3 triangle_refine(KernelGlobals kg,
-                                         ccl_private ShaderData *sd,
-                                         float3 P,
-                                         float3 D,
-                                         float t,
-                                         const int isect_object,
-                                         const int isect_prim)
+ccl_device_inline float3 triangle_point_from_uv(KernelGlobals kg,
+                                                ccl_private ShaderData *sd,
+                                                const int isect_object,
+                                                const int isect_prim,
+                                                const float u,
+                                                const float v)
 {
-#ifdef __INTERSECTION_REFINE__
-  if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
-    if (UNLIKELY(t == 0.0f)) {
-      return P;
-    }
-    const Transform tfm = object_get_inverse_transform(kg, sd);
-
-    P = transform_point(&tfm, P);
-    D = transform_direction(&tfm, D * t);
-    D = normalize_len(D, &t);
-  }
-
-  P = P + D * t;
-
   const uint tri_vindex = kernel_tex_fetch(__tri_vindex, isect_prim).w;
   const packed_float3 tri_a = kernel_tex_fetch(__tri_verts, tri_vindex + 0),
                       tri_b = kernel_tex_fetch(__tri_verts, tri_vindex + 1),
                       tri_c = kernel_tex_fetch(__tri_verts, tri_vindex + 2);
-  float3 edge1 = make_float3(tri_a.x - tri_c.x, tri_a.y - tri_c.y, tri_a.z - tri_c.z);
-  float3 edge2 = make_float3(tri_b.x - tri_c.x, tri_b.y - tri_c.y, tri_b.z - tri_c.z);
-  float3 tvec = make_float3(P.x - tri_c.x, P.y - tri_c.y, P.z - tri_c.z);
-  float3 qvec = cross(tvec, edge1);
-  float3 pvec = cross(D, edge2);
-  float det = dot(edge1, pvec);
-  if (det != 0.0f) {
-    /* If determinant is zero it means ray lies in the plane of
-     * the triangle. It is possible in theory due to watertight
-     * nature of triangle intersection. For such cases we simply
-     * don't refine intersection hoping it'll go all fine.
-     */
-    float rt = dot(edge2, qvec) / det;
-    P = P + D * rt;
-  }
+  float w = 1.0f - u - v;
+
+  float3 P = u * tri_a + v * tri_b + w * tri_c;
 
   if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
     const Transform tfm = object_get_transform(kg, sd);
@@ -201,65 +166,6 @@ ccl_device_inline float3 triangle_refine(KernelGlobals kg,
   }
 
   return P;
-#else
-  return P + D * t;
-#endif
-}
-
-/* Same as above, except that t is assumed to be in object space for
- * instancing.
- */
-ccl_device_inline float3 triangle_refine_local(KernelGlobals kg,
-                                               ccl_private ShaderData *sd,
-                                               float3 P,
-                                               float3 D,
-                                               float t,
-                                               const int isect_object,
-                                               const int isect_prim)
-{
-#if defined(__KERNEL_GPU_RAYTRACING__)
-  /* t is always in world space with OptiX and MetalRT. */
-  return triangle_refine(kg, sd, P, D, t, isect_object, isect_prim);
-#else
-  if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
-    const Transform tfm = object_get_inverse_transform(kg, sd);
-
-    P = transform_point(&tfm, P);
-    D = transform_direction(&tfm, D);
-    D = normalize(D);
-  }
-
-  P = P + D * t;
-
-#  ifdef __INTERSECTION_REFINE__
-  const uint tri_vindex = kernel_tex_fetch(__tri_vindex, isect_prim).w;
-  const packed_float3 tri_a = kernel_tex_fetch(__tri_verts, tri_vindex + 0),
-                      tri_b = kernel_tex_fetch(__tri_verts, tri_vindex + 1),
-                      tri_c = kernel_tex_fetch(__tri_verts, tri_vindex + 2);
-  float3 edge1 = make_float3(tri_a.x - tri_c.x, tri_a.y - tri_c.y, tri_a.z - tri_c.z);
-  float3 edge2 = make_float3(tri_b.x - tri_c.x, tri_b.y - tri_c.y, tri_b.z - tri_c.z);
-  float3 tvec = make_float3(P.x - tri_c.x, P.y - tri_c.y, P.z - tri_c.z);
-  float3 qvec = cross(tvec, edge1);
-  float3 pvec = cross(D, edge2);
-  float det = dot(edge1, pvec);
-  if (det != 0.0f) {
-    /* If determinant is zero it means ray lies in the plane of
-     * the triangle. It is possible in theory due to watertight
-     * nature of triangle intersection. For such cases we simply
-     * don't refine intersection hoping it'll go all fine.
-     */
-    float rt = dot(edge2, qvec) / det;
-    P = P + D * rt;
-  }
-#  endif /* __INTERSECTION_REFINE__ */
-
-  if (!(sd->object_flag & SD_OBJECT_TRANSFORM_APPLIED)) {
-    const Transform tfm = object_get_transform(kg, sd);
-    P = transform_point(&tfm, P);
-  }
-
-  return P;
-#endif
 }
 
 CCL_NAMESPACE_END

intern/cycles/kernel/integrator/intersect_closest.h

@@ -328,6 +328,12 @@ ccl_device void integrator_intersect_closest(KernelGlobals kg,
 
   /* Scene Intersection. */
   Intersection isect ccl_optional_struct_init;
+  isect.object = OBJECT_NONE;
+  isect.prim = PRIM_NONE;
+  ray.self.object = last_isect_object;
+  ray.self.prim = last_isect_prim;
+  ray.self.light_object = OBJECT_NONE;
+  ray.self.light_prim = PRIM_NONE;
   bool hit = scene_intersect(kg, &ray, visibility, &isect);
 
   /* TODO: remove this and do it in the various intersection functions instead. */

intern/cycles/kernel/integrator/intersect_shadow.h

@@ -156,7 +156,10 @@ ccl_device void integrator_intersect_shadow(KernelGlobals kg, IntegratorShadowSt
   /* Read ray from integrator state into local memory. */
   Ray ray ccl_optional_struct_init;
   integrator_state_read_shadow_ray(kg, state, &ray);
-
+  ray.self.object = INTEGRATOR_STATE_ARRAY(state, shadow_isect, 0, object);
+  ray.self.prim = INTEGRATOR_STATE_ARRAY(state, shadow_isect, 0, prim);
+  ray.self.light_object = INTEGRATOR_STATE_ARRAY(state, shadow_isect, 1, object);
+  ray.self.light_prim = INTEGRATOR_STATE_ARRAY(state, shadow_isect, 1, prim);
   /* Compute visibility. */
   const uint visibility = integrate_intersect_shadow_visibility(kg, state);
 

intern/cycles/kernel/integrator/intersect_volume_stack.h

@@ -38,7 +38,10 @@ ccl_device void integrator_volume_stack_update_for_subsurface(KernelGlobals kg,
   Ray volume_ray ccl_optional_struct_init;
   volume_ray.P = from_P;
   volume_ray.D = normalize_len(to_P - from_P, &volume_ray.t);
-
+  volume_ray.self.object = INTEGRATOR_STATE(state, isect, object);
+  volume_ray.self.prim = INTEGRATOR_STATE(state, isect, prim);
+  volume_ray.self.light_object = OBJECT_NONE;
+  volume_ray.self.light_prim = PRIM_NONE;
   /* Store to avoid global fetches on every intersection step. */
   const uint volume_stack_size = kernel_data.volume_stack_size;
 
@@ -68,7 +71,7 @@ ccl_device void integrator_volume_stack_update_for_subsurface(KernelGlobals kg,
     volume_stack_enter_exit(kg, state, stack_sd);
 
     /* Move ray forward. */
-    volume_ray.P = ray_offset(stack_sd->P, -stack_sd->Ng);
+    volume_ray.P = stack_sd->P;
     if (volume_ray.t != FLT_MAX) {
       volume_ray.D = normalize_len(to_P - volume_ray.P, &volume_ray.t);
     }
@@ -91,6 +94,10 @@ ccl_device void integrator_volume_stack_init(KernelGlobals kg, IntegratorState s
    * fewest hits. */
   volume_ray.D = make_float3(0.0f, 0.0f, 1.0f);
   volume_ray.t = FLT_MAX;
+  volume_ray.self.object = OBJECT_NONE;
+  volume_ray.self.prim = PRIM_NONE;
+  volume_ray.self.light_object = OBJECT_NONE;
+  volume_ray.self.light_prim = PRIM_NONE;
 
   int stack_index = 0, enclosed_index = 0;
 
@@ -203,7 +210,7 @@ ccl_device void integrator_volume_stack_init(KernelGlobals kg, IntegratorState s
     }
 
     /* Move ray forward. */
-    volume_ray.P = ray_offset(stack_sd->P, -stack_sd->Ng);
+    volume_ray.P = stack_sd->P;
     ++step;
   }
 #endif

intern/cycles/kernel/integrator/shade_light.h

@@ -37,8 +37,9 @@ ccl_device_inline void integrate_light(KernelGlobals kg,
 
   /* Advance ray beyond light. */
   /* TODO: can we make this more numerically robust to avoid reintersecting the
-   * same light in some cases? */
-  const float3 new_ray_P = ray_offset(ray_P + ray_D * isect.t, ray_D);
+   * same light in some cases? Ray should not intersect surface anymore as the
+   * object and prim ids will prevent self intersection. */
+  const float3 new_ray_P = ray_P + ray_D * isect.t;
   INTEGRATOR_STATE_WRITE(state, ray, P) = new_ray_P;
   INTEGRATOR_STATE_WRITE(state, ray, t) -= isect.t;
 
@@ -46,7 +47,7 @@ ccl_device_inline void integrate_light(KernelGlobals kg,
   const float mis_ray_t = INTEGRATOR_STATE(state, path, mis_ray_t);
   ray_P -= ray_D * mis_ray_t;
   isect.t += mis_ray_t;
-  INTEGRATOR_STATE_WRITE(state, path, mis_ray_t) = mis_ray_t + isect.t;
+  INTEGRATOR_STATE_WRITE(state, path, mis_ray_t) = isect.t;
 
   LightSample ls ccl_optional_struct_init;
   const bool use_light_sample = light_sample_from_intersection(kg, &isect, ray_P, ray_D, &ls);

intern/cycles/kernel/integrator/shade_shadow.h

@@ -83,7 +83,10 @@ ccl_device_inline void integrate_transparent_volume_shadow(KernelGlobals kg,
   /* Setup shader data. */
   Ray ray ccl_optional_struct_init;
   integrator_state_read_shadow_ray(kg, state, &ray);
-
+  ray.self.object = OBJECT_NONE;
+  ray.self.prim = PRIM_NONE;
+  ray.self.light_object = OBJECT_NONE;
+  ray.self.light_prim = PRIM_NONE;
   /* Modify ray position and length to match current segment. */
   const float start_t = (hit == 0) ? 0.0f :
                                      INTEGRATOR_STATE_ARRAY(state, shadow_isect, hit - 1, t);
@@ -149,7 +152,7 @@ ccl_device_inline bool integrate_transparent_shadow(KernelGlobals kg,
     const float last_hit_t = INTEGRATOR_STATE_ARRAY(state, shadow_isect, num_recorded_hits - 1, t);
     const float3 ray_P = INTEGRATOR_STATE(state, shadow_ray, P);
     const float3 ray_D = INTEGRATOR_STATE(state, shadow_ray, D);
-    INTEGRATOR_STATE_WRITE(state, shadow_ray, P) = ray_offset(ray_P + last_hit_t * ray_D, ray_D);
+    INTEGRATOR_STATE_WRITE(state, shadow_ray, P) = ray_P + last_hit_t * ray_D;
     INTEGRATOR_STATE_WRITE(state, shadow_ray, t) -= last_hit_t;
   }
 

intern/cycles/kernel/integrator/shade_surface.h

@@ -182,6 +182,11 @@ ccl_device_forceinline void integrate_surface_direct_light(KernelGlobals kg,
 
   /* Write shadow ray and associated state to global memory. */
   integrator_state_write_shadow_ray(kg, shadow_state, &ray);
+  // Save memory by storing the light and object indices in the shadow_isect
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 0, object) = ray.self.object;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 0, prim) = ray.self.prim;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 1, object) = ray.self.light_object;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 1, prim) = ray.self.light_prim;
 
   /* Copy state from main path to shadow path. */
   const uint16_t bounce = INTEGRATOR_STATE(state, path, bounce);
@@ -266,13 +271,11 @@ ccl_device_forceinline int integrate_surface_bsdf_bssrdf_bounce(
   }
 
   /* Setup ray. Note that clipping works through transparent bounces. */
-  INTEGRATOR_STATE_WRITE(state, ray, P) = ray_offset(sd->P,
-                                                     (label & LABEL_TRANSMIT) ? -sd->Ng : sd->Ng);
+  INTEGRATOR_STATE_WRITE(state, ray, P) = sd->P;
   INTEGRATOR_STATE_WRITE(state, ray, D) = normalize(bsdf_omega_in);
   INTEGRATOR_STATE_WRITE(state, ray, t) = (label & LABEL_TRANSPARENT) ?
                                               INTEGRATOR_STATE(state, ray, t) - sd->ray_length :
                                               FLT_MAX;
-
 #ifdef __RAY_DIFFERENTIALS__
   INTEGRATOR_STATE_WRITE(state, ray, dP) = differential_make_compact(sd->dP);
   INTEGRATOR_STATE_WRITE(state, ray, dD) = differential_make_compact(bsdf_domega_in);
@@ -316,7 +319,7 @@ ccl_device_forceinline bool integrate_surface_volume_only_bounce(IntegratorState
   }
 
   /* Setup ray position, direction stays unchanged. */
-  INTEGRATOR_STATE_WRITE(state, ray, P) = ray_offset(sd->P, -sd->Ng);
+  INTEGRATOR_STATE_WRITE(state, ray, P) = sd->P;
 
   /* Clipping works through transparent. */
   INTEGRATOR_STATE_WRITE(state, ray, t) -= sd->ray_length;
@@ -360,10 +363,14 @@ ccl_device_forceinline void integrate_surface_ao(KernelGlobals kg,
   }
 
   Ray ray ccl_optional_struct_init;
-  ray.P = ray_offset(sd->P, sd->Ng);
+  ray.P = sd->P;
   ray.D = ao_D;
   ray.t = kernel_data.integrator.ao_bounces_distance;
   ray.time = sd->time;
+  ray.self.object = sd->object;
+  ray.self.prim = sd->prim;
+  ray.self.light_object = OBJECT_NONE;
+  ray.self.light_prim = PRIM_NONE;
   ray.dP = differential_zero_compact();
   ray.dD = differential_zero_compact();
 
@@ -375,6 +382,10 @@ ccl_device_forceinline void integrate_surface_ao(KernelGlobals kg,
 
   /* Write shadow ray and associated state to global memory. */
   integrator_state_write_shadow_ray(kg, shadow_state, &ray);
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 0, object) = ray.self.object;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 0, prim) = ray.self.prim;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 1, object) = ray.self.light_object;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 1, prim) = ray.self.light_prim;
 
   /* Copy state from main path to shadow path. */
   const uint16_t bounce = INTEGRATOR_STATE(state, path, bounce);

intern/cycles/kernel/integrator/shade_volume.h

@@ -791,6 +791,10 @@ ccl_device_forceinline void integrate_volume_direct_light(
 
   /* Write shadow ray and associated state to global memory. */
   integrator_state_write_shadow_ray(kg, shadow_state, &ray);
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 0, object) = ray.self.object;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 0, prim) = ray.self.prim;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 1, object) = ray.self.light_object;
+  INTEGRATOR_STATE_ARRAY_WRITE(shadow_state, shadow_isect, 1, prim) = ray.self.light_prim;
 
   /* Copy state from main path to shadow path. */
   const uint16_t bounce = INTEGRATOR_STATE(state, path, bounce);
@@ -873,11 +877,13 @@ ccl_device_forceinline bool integrate_volume_phase_scatter(
   INTEGRATOR_STATE_WRITE(state, ray, P) = sd->P;
   INTEGRATOR_STATE_WRITE(state, ray, D) = normalize(phase_omega_in);
   INTEGRATOR_STATE_WRITE(state, ray, t) = FLT_MAX;
-
 #  ifdef __RAY_DIFFERENTIALS__
   INTEGRATOR_STATE_WRITE(state, ray, dP) = differential_make_compact(sd->dP);
   INTEGRATOR_STATE_WRITE(state, ray, dD) = differential_make_compact(phase_domega_in);
 #  endif
+  // Save memory by storing last hit prim and object in isect
+  INTEGRATOR_STATE_WRITE(state, isect, prim) = sd->prim;
+  INTEGRATOR_STATE_WRITE(state, isect, object) = sd->object;
 
   /* Update throughput. */
   const float3 throughput = INTEGRATOR_STATE(state, path, throughput);

intern/cycles/kernel/integrator/shadow_state_template.h

@@ -61,6 +61,7 @@ KERNEL_STRUCT_MEMBER(shadow_ray, packed_float3, D, KERNEL_FEATURE_PATH_TRACING)
 KERNEL_STRUCT_MEMBER(shadow_ray, float, t, KERNEL_FEATURE_PATH_TRACING)
 KERNEL_STRUCT_MEMBER(shadow_ray, float, time, KERNEL_FEATURE_PATH_TRACING)
 KERNEL_STRUCT_MEMBER(shadow_ray, float, dP, KERNEL_FEATURE_PATH_TRACING)
+KERNEL_STRUCT_MEMBER(shadow_ray, int, object, KERNEL_FEATURE_PATH_TRACING)
 KERNEL_STRUCT_END(shadow_ray)
 
 /*********************** Shadow Intersection result **************************/

intern/cycles/kernel/integrator/subsurface.h

@@ -57,7 +57,6 @@ ccl_device int subsurface_bounce(KernelGlobals kg,
 
   /* Pass along object info, reusing isect to save memory. */
   INTEGRATOR_STATE_WRITE(state, subsurface, Ng) = sd->Ng;
-  INTEGRATOR_STATE_WRITE(state, isect, object) = sd->object;
 
   uint32_t path_flag = (INTEGRATOR_STATE(state, path, flag) & ~PATH_RAY_CAMERA) |
                        ((sc->type == CLOSURE_BSSRDF_BURLEY_ID) ? PATH_RAY_SUBSURFACE_DISK :
@@ -165,10 +164,8 @@ ccl_device_inline bool subsurface_scatter(KernelGlobals kg, IntegratorState stat
 
     if (object_flag & SD_OBJECT_INTERSECTS_VOLUME) {
       float3 P = INTEGRATOR_STATE(state, ray, P);
-      const float3 Ng = INTEGRATOR_STATE(state, subsurface, Ng);
-      const float3 offset_P = ray_offset(P, -Ng);
 
-      integrator_volume_stack_update_for_subsurface(kg, state, offset_P, ray.P);
+      integrator_volume_stack_update_for_subsurface(kg, state, P, ray.P);
     }
   }
 #  endif /* __VOLUME__ */

intern/cycles/kernel/integrator/subsurface_disk.h

@@ -99,6 +99,10 @@ ccl_device_inline bool subsurface_disk(KernelGlobals kg,
   ray.dP = ray_dP;
   ray.dD = differential_zero_compact();
   ray.time = time;
+  ray.self.object = OBJECT_NONE;
+  ray.self.prim = PRIM_NONE;
+  ray.self.light_object = OBJECT_NONE;
+  ray.self.light_prim = OBJECT_NONE;
 
   /* Intersect with the same object. if multiple intersections are found it
    * will use at most BSSRDF_MAX_HITS hits, a random subset of all hits. */

intern/cycles/kernel/integrator/subsurface_random_walk.h

@@ -195,6 +195,7 @@ ccl_device_inline bool subsurface_random_walk(KernelGlobals kg,
   const float time = INTEGRATOR_STATE(state, ray, time);
   const float3 Ng = INTEGRATOR_STATE(state, subsurface, Ng);
   const int object = INTEGRATOR_STATE(state, isect, object);
+  const int prim = INTEGRATOR_STATE(state, isect, prim);
 
   /* Sample diffuse surface scatter into the object. */
   float3 D;
@@ -205,12 +206,16 @@ ccl_device_inline bool subsurface_random_walk(KernelGlobals kg,
   }
 
   /* Setup ray. */
-  ray.P = ray_offset(P, -Ng);
+  ray.P = P;
   ray.D = D;
   ray.t = FLT_MAX;
   ray.time = time;
   ray.dP = ray_dP;
   ray.dD = differential_zero_compact();
+  ray.self.object = object;
+  ray.self.prim = prim;
+  ray.self.light_object = OBJECT_NONE;
+  ray.self.light_prim = PRIM_NONE;
 
 #ifndef __KERNEL_GPU_RAYTRACING__
   /* Compute or fetch object transforms. */
@@ -377,7 +382,15 @@ ccl_device_inline bool subsurface_random_walk(KernelGlobals kg,
      * If yes, we will later use backwards guided sampling in order to have a decent
      * chance of connecting to it.
      * TODO: Maybe use less than 10 times the mean free path? */
-    ray.t = (bounce == 0) ? max(t, 10.0f / (min3(sigma_t))) : t;
+    if (bounce == 0) {
+      ray.t = max(t, 10.0f / (min3(sigma_t)));
+    }
+    else {
+      ray.t = t;
+      /* After the first bounce the object can intersect the same surface again */
+      ray.self.object = OBJECT_NONE;
+      ray.self.prim = PRIM_NONE;
+    }
     scene_intersect_local(kg, &ray, &ss_isect, object, NULL, 1);
     hit = (ss_isect.num_hits > 0);
 
@@ -408,13 +421,6 @@ ccl_device_inline bool subsurface_random_walk(KernelGlobals kg,
     if (hit) {
       t = ray.t;
     }
-    else if (bounce == 0) {
-      /* Restore original position if nothing was hit after the first bounce,
-       * without the ray_offset() that was added to avoid self-intersection.
-       * Otherwise if that offset is relatively large compared to the scattering
-       * radius, we never go back up high enough to exit the surface. */
-      ray.P = P;
-    }
 
     /* Advance to new scatter location. */
     ray.P += t * ray.D;

intern/cycles/kernel/light/light.h

@@ -418,8 +418,8 @@ ccl_device bool light_sample_from_intersection(KernelGlobals kg,
   LightType type = (LightType)klight->type;
   ls->type = type;
   ls->shader = klight->shader_id;
-  ls->object = PRIM_NONE;
-  ls->prim = PRIM_NONE;
+  ls->object = isect->object;
+  ls->prim = isect->prim;
   ls->lamp = lamp;
   /* todo: missing texture coordinates */
   ls->t = isect->t;

intern/cycles/kernel/light/sample.h

@@ -198,7 +198,7 @@ ccl_device_inline float3 shadow_ray_offset(KernelGlobals kg,
   float NL = dot(sd->N, L);
   bool transmit = (NL < 0.0f);
   float3 Ng = (transmit ? -sd->Ng : sd->Ng);
-  float3 P = ray_offset(sd->P, Ng);
+  float3 P = sd->P;
 
   if ((sd->type & PRIMITIVE_TRIANGLE) && (sd->shader & SHADER_SMOOTH_NORMAL)) {
     const float offset_cutoff =
@@ -243,7 +243,7 @@ ccl_device_inline void shadow_ray_setup(ccl_private const ShaderData *ccl_restri
     }
     else {
       /* other lights, avoid self-intersection */
-      ray->D = ray_offset(ls->P, ls->Ng) - P;
+      ray->D = ls->P - P;
       ray->D = normalize_len(ray->D, &ray->t);
     }
   }
@@ -257,6 +257,12 @@ ccl_device_inline void shadow_ray_setup(ccl_private const ShaderData *ccl_restri
   ray->dP = differential_make_compact(sd->dP);
   ray->dD = differential_zero_compact();
   ray->time = sd->time;
+
+  /* Fill in intersection surface and light details. */
+  ray->self.prim = sd->prim;
+  ray->self.object = sd->object;
+  ray->self.light_prim = ls->prim;
+  ray->self.light_object = ls->object;
 }
 
 /* Create shadow ray towards light sample. */

intern/cycles/kernel/svm/ao.h

@@ -70,10 +70,14 @@ ccl_device float svm_ao(
 
     /* Create ray. */
     Ray ray;
-    ray.P = ray_offset(sd->P, N);
+    ray.P = sd->P;
     ray.D = D.x * T + D.y * B + D.z * N;
     ray.t = max_dist;
     ray.time = sd->time;
+    ray.self.object = sd->object;
+    ray.self.prim = sd->prim;
+    ray.self.light_object = OBJECT_NONE;
+    ray.self.light_prim = PRIM_NONE;
     ray.dP = differential_zero_compact();
     ray.dD = differential_zero_compact();
 

intern/cycles/kernel/svm/bevel.h

@@ -196,6 +196,10 @@ ccl_device float3 svm_bevel(
     ray.dP = differential_zero_compact();
     ray.dD = differential_zero_compact();
     ray.time = sd->time;
+    ray.self.object = OBJECT_NONE;
+    ray.self.prim = PRIM_NONE;
+    ray.self.light_object = OBJECT_NONE;
+    ray.self.light_prim = PRIM_NONE;
 
     /* Intersect with the same object. if multiple intersections are found it
      * will use at most LOCAL_MAX_HITS hits, a random subset of all hits. */
@@ -207,15 +211,24 @@ ccl_device float3 svm_bevel(
       /* Quickly retrieve P and Ng without setting up ShaderData. */
       float3 hit_P;
       if (sd->type == PRIMITIVE_TRIANGLE) {
-        hit_P = triangle_refine_local(
-            kg, sd, ray.P, ray.D, ray.t, isect.hits[hit].object, isect.hits[hit].prim);
+        hit_P = triangle_point_from_uv(kg,
+                                       sd,
+                                       isect.hits[hit].object,
+                                       isect.hits[hit].prim,
+                                       isect.hits[hit].u,
+                                       isect.hits[hit].v);
       }
 #  ifdef __OBJECT_MOTION__
       else if (sd->type == PRIMITIVE_MOTION_TRIANGLE) {
         float3 verts[3];
         motion_triangle_vertices(kg, sd->object, isect.hits[hit].prim, sd->time, verts);
-        hit_P = motion_triangle_refine_local(
-            kg, sd, ray.P, ray.D, ray.t, isect.hits[hit].object, isect.hits[hit].prim, verts);
+        hit_P = motion_triangle_point_from_uv(kg,
+                                              sd,
+                                              isect.hits[hit].object,
+                                              isect.hits[hit].prim,
+                                              isect.hits[hit].u,
+                                              isect.hits[hit].v,
+                                              verts);
       }
 #  endif /* __OBJECT_MOTION__ */
 

intern/cycles/kernel/types.h

@@ -512,12 +512,21 @@ typedef struct differential {
 
 /* Ray */
 
+typedef struct RaySelfPrimitives {
+  int prim;         /* Primitive the ray is starting from */
+  int object;       /* Instance prim is a part of */
+  int light_prim;   /* Light primitive */
+  int light_object; /* Light object */
+} RaySelfPrimitives;
+
 typedef struct Ray {
   float3 P;   /* origin */
   float3 D;   /* direction */
   float t;    /* length of the ray */
   float time; /* time (for motion blur) */
 
+  RaySelfPrimitives self;
+
 #ifdef __RAY_DIFFERENTIALS__
   float dP;
   float dD;

intern/ghost/test/CMakeLists.txt

@@ -292,7 +292,7 @@ target_link_libraries(multitest_c
   guardedalloc_lib
   wcwidth_lib
   ${OPENGL_gl_LIBRARY}
-  ${FREETYPE_LIBRARIES}
+  ${FREETYPE_LIBRARIES} ${BROTLI_LIBRARIES}
   ${ZLIB_LIBRARIES}
   ${CMAKE_DL_LIBS}
   ${PLATFORM_LINKLIBS}

release/scripts/startup/bl_ui/space_filebrowser.py

@@ -761,6 +761,15 @@ class ASSETBROWSER_PT_metadata_preview(asset_utils.AssetMetaDataPanel, Panel):
         col.operator("ed.lib_id_load_custom_preview", icon='FILEBROWSER', text="")
         col.separator()
         col.operator("ed.lib_id_generate_preview", icon='FILE_REFRESH', text="")
+        col.menu("ASSETBROWSER_MT_metadata_preview_menu", icon='DOWNARROW_HLT', text="")
+
+
+class ASSETBROWSER_MT_metadata_preview_menu(bpy.types.Menu):
+    bl_label = "Preview"
+
+    def draw(self, context):
+        layout = self.layout
+        layout.operator("ed.lib_id_generate_preview_from_object", text="Render Active Object")
 
 
 class ASSETBROWSER_PT_metadata_tags(asset_utils.AssetMetaDataPanel, Panel):
@@ -840,6 +849,7 @@ classes = (
     ASSETBROWSER_MT_view,
     ASSETBROWSER_MT_select,
     ASSETBROWSER_MT_edit,
+    ASSETBROWSER_MT_metadata_preview_menu,
     ASSETBROWSER_PT_metadata,
     ASSETBROWSER_PT_metadata_preview,
     ASSETBROWSER_PT_metadata_tags,

source/blender/blenfont/CMakeLists.txt

@@ -54,7 +54,7 @@ set(LIB
   bf_gpu
   bf_intern_guardedalloc
 
-  ${FREETYPE_LIBRARIES}
+  ${FREETYPE_LIBRARIES} ${BROTLI_LIBRARIES}
 )
 
 if(WIN32)

source/blender/blenkernel/BKE_lib_remap.h

@@ -38,6 +38,9 @@
 extern "C" {
 #endif
 
+struct ID;
+struct IDRemapper;
+
 /* BKE_libblock_free, delete are declared in BKE_lib_id.h for convenience. */
 
 /* Also IDRemap->flag. */
@@ -97,6 +100,19 @@ enum {
   ID_REMAP_FORCE_OBDATA_IN_EDITMODE = 1 << 9,
 };
 
+/**
+ * Replace all references in given Main using the given \a mappings
+ *
+ * \note Is preferred over BKE_libblock_remap_locked due to performance.
+ */
+void BKE_libblock_remap_multiple_locked(struct Main *bmain,
+                                        const struct IDRemapper *mappings,
+                                        const short remap_flags);
+
+void BKE_libblock_remap_multiple(struct Main *bmain,
+                                 const struct IDRemapper *mappings,
+                                 const short remap_flags);
+
 /**
  * Replace all references in given Main to \a old_id by \a new_id
  * (if \a new_id is NULL, it unlinks \a old_id).
@@ -146,12 +162,61 @@ void BKE_libblock_relink_to_newid(struct Main *bmain, struct ID *id, int remap_f
     ATTR_NONNULL();
 
 typedef void (*BKE_library_free_notifier_reference_cb)(const void *);
-typedef void (*BKE_library_remap_editor_id_reference_cb)(struct ID *, struct ID *);
+typedef void (*BKE_library_remap_editor_id_reference_cb)(const struct IDRemapper *mappings);
 
 void BKE_library_callback_free_notifier_reference_set(BKE_library_free_notifier_reference_cb func);
 void BKE_library_callback_remap_editor_id_reference_set(
     BKE_library_remap_editor_id_reference_cb func);
 
+/* IDRemapper */
+struct IDRemapper;
+typedef enum IDRemapperApplyResult {
+  /** No remapping rules available for the source. */
+  ID_REMAP_RESULT_SOURCE_UNAVAILABLE,
+  /** Source isn't mappable (e.g. NULL). */
+  ID_REMAP_RESULT_SOURCE_NOT_MAPPABLE,
+  /** Source has been remapped to a new pointer. */
+  ID_REMAP_RESULT_SOURCE_REMAPPED,
+  /** Source has been set to NULL. */
+  ID_REMAP_RESULT_SOURCE_UNASSIGNED,
+} IDRemapperApplyResult;
+
+typedef enum IDRemapperApplyOptions {
+  ID_REMAP_APPLY_UPDATE_REFCOUNT = (1 << 0),
+  ID_REMAP_APPLY_ENSURE_REAL = (1 << 1),
+
+  ID_REMAP_APPLY_DEFAULT = 0,
+} IDRemapperApplyOptions;
+
+typedef void (*IDRemapperIterFunction)(struct ID *old_id, struct ID *new_id, void *user_data);
+
+/**
+ * Create a new ID Remapper.
+ *
+ * An ID remapper stores multiple remapping rules.
+ */
+struct IDRemapper *BKE_id_remapper_create(void);
+
+void BKE_id_remapper_clear(struct IDRemapper *id_remapper);
+bool BKE_id_remapper_is_empty(const struct IDRemapper *id_remapper);
+/** Free the given ID Remapper. */
+void BKE_id_remapper_free(struct IDRemapper *id_remapper);
+/** Add a new remapping. */
+void BKE_id_remapper_add(struct IDRemapper *id_remapper, struct ID *old_id, struct ID *new_id);
+
+/**
+ * Apply a remapping.
+ *
+ * Update the id pointer stored in the given r_id_ptr if a remapping rule exists.
+ */
+IDRemapperApplyResult BKE_id_remapper_apply(const struct IDRemapper *id_remapper,
+                                            struct ID **r_id_ptr,
+                                            IDRemapperApplyOptions options);
+bool BKE_id_remapper_has_mapping_for(const struct IDRemapper *id_remapper, uint64_t type_filter);
+void BKE_id_remapper_iter(const struct IDRemapper *id_remapper,
+                          IDRemapperIterFunction func,
+                          void *user_data);
+
 #ifdef __cplusplus
 }
 #endif

source/blender/blenkernel/BKE_screen.h

@@ -38,6 +38,7 @@ struct BlendLibReader;
 struct BlendWriter;
 struct Header;
 struct ID;
+struct IDRemapper;
 struct LibraryForeachIDData;
 struct ListBase;
 struct Menu;
@@ -117,10 +118,7 @@ typedef struct SpaceType {
   bContextDataCallback context;
 
   /* Used when we want to replace an ID by another (or NULL). */
-  void (*id_remap)(struct ScrArea *area,
-                   struct SpaceLink *sl,
-                   struct ID *old_id,
-                   struct ID *new_id);
+  void (*id_remap)(struct ScrArea *area, struct SpaceLink *sl, const struct IDRemapper *mappings);
 
   int (*space_subtype_get)(struct ScrArea *area);
   void (*space_subtype_set)(struct ScrArea *area, int value);

source/blender/blenkernel/CMakeLists.txt

@@ -179,6 +179,7 @@ set(SRC
   intern/lib_id.c
   intern/lib_id_delete.c
   intern/lib_id_eval.c
+  intern/lib_id_remapper.cc
   intern/lib_override.c
   intern/lib_query.c
   intern/lib_remap.c
@@ -522,7 +523,7 @@ set(LIB
   bf_simulation
 
   # For `vfontdata_freetype.c`.
-  ${FREETYPE_LIBRARIES}
+  ${FREETYPE_LIBRARIES} ${BROTLI_LIBRARIES}
 )
 
 if(WITH_BINRELOC)
@@ -823,6 +824,7 @@ if(WITH_GTESTS)
     intern/idprop_serialize_test.cc
     intern/lattice_deform_test.cc
     intern/layer_test.cc
+    intern/lib_id_remapper_test.cc
     intern/lib_id_test.cc
     intern/lib_remap_test.cc
     intern/tracking_test.cc

source/blender/blenkernel/intern/lib_id_delete.c

@@ -154,7 +154,10 @@ void BKE_id_free_ex(Main *bmain, void *idv, int flag, const bool use_flag_from_i
     }
 
     if (remap_editor_id_reference_cb) {
-      remap_editor_id_reference_cb(id, NULL);
+      struct IDRemapper *remapper = BKE_id_remapper_create();
+      BKE_id_remapper_add(remapper, id, NULL);
+      remap_editor_id_reference_cb(remapper);
+      BKE_id_remapper_free(remapper);
     }
   }
 
@@ -292,32 +295,40 @@ static size_t id_delete(Main *bmain, const bool do_tagged_deletion)
      * Note that we go forward here, since we want to check dependencies before users
      * (e.g. meshes before objects).
      * Avoids to have to loop twice. */
+    struct IDRemapper *remapper = BKE_id_remapper_create();
     for (i = 0; i < base_count; i++) {
       ListBase *lb = lbarray[i];
       ID *id, *id_next;
+      BKE_id_remapper_clear(remapper);
 
       for (id = lb->first; id; id = id_next) {
         id_next = id->next;
         /* NOTE: in case we delete a library, we also delete all its datablocks! */
         if ((id->tag & tag) || (id->lib != NULL && (id->lib->id.tag & tag))) {
           id->tag |= tag;
-
-          /* Will tag 'never NULL' users of this ID too.
-           * Note that we cannot use BKE_libblock_unlink() here, since it would ignore indirect
-           * (and proxy!) links, this can lead to nasty crashing here in second,
-           * actual deleting loop.
-           * Also, this will also flag users of deleted data that cannot be unlinked
-           * (object using deleted obdata, etc.), so that they also get deleted. */
-          BKE_libblock_remap_locked(bmain,
-                                    id,
-                                    NULL,
-                                    (ID_REMAP_FLAG_NEVER_NULL_USAGE |
-                                     ID_REMAP_FORCE_NEVER_NULL_USAGE |
-                                     ID_REMAP_FORCE_INTERNAL_RUNTIME_POINTERS));
+          BKE_id_remapper_add(remapper, id, NULL);
         }
       }
+
+      if (BKE_id_remapper_is_empty(remapper)) {
+        continue;
+      }
+
+      /* Will tag 'never NULL' users of this ID too.
+       * Note that we cannot use BKE_libblock_unlink() here, since it would ignore indirect
+       * (and proxy!) links, this can lead to nasty crashing here in second,
+       * actual deleting loop.
+       * Also, this will also flag users of deleted data that cannot be unlinked
+       * (object using deleted obdata, etc.), so that they also get deleted. */
+      BKE_libblock_remap_multiple_locked(bmain,
+                                         remapper,
+                                         (ID_REMAP_FLAG_NEVER_NULL_USAGE |
+                                          ID_REMAP_FORCE_NEVER_NULL_USAGE |
+                                          ID_REMAP_FORCE_INTERNAL_RUNTIME_POINTERS));
     }
+    BKE_id_remapper_free(remapper);
   }
+
   BKE_main_unlock(bmain);
 
   /* In usual reversed order, such that all usage of a given ID, even 'never NULL' ones,

source/blender/blenkernel/intern/lib_id_remapper.cc

@@ -0,0 +1,175 @@
+/*
+ * 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) 2022 by Blender Foundation.
+ */
+
+#include "DNA_ID.h"
+
+#include "BKE_idtype.h"
+#include "BKE_lib_id.h"
+#include "BKE_lib_remap.h"
+
+#include "MEM_guardedalloc.h"
+
+#include "BLI_map.hh"
+
+using IDTypeFilter = uint64_t;
+
+namespace blender::bke::id::remapper {
+struct IDRemapper {
+ private:
+  Map<ID *, ID *> mappings;
+  IDTypeFilter source_types = 0;
+
+ public:
+  void clear()
+  {
+    mappings.clear();
+    source_types = 0;
+  }
+
+  bool is_empty() const
+  {
+    return mappings.is_empty();
+  }
+
+  void add(ID *old_id, ID *new_id)
+  {
+    BLI_assert(old_id != nullptr);
+    BLI_assert(new_id == nullptr || (GS(old_id->name) == GS(new_id->name)));
+    mappings.add(old_id, new_id);
+    source_types |= BKE_idtype_idcode_to_idfilter(GS(old_id->name));
+  }
+
+  bool contains_mappings_for_any(IDTypeFilter filter) const
+  {
+    return (source_types & filter) != 0;
+  }
+
+  IDRemapperApplyResult apply(ID **r_id_ptr, IDRemapperApplyOptions options) const
+  {
+    BLI_assert(r_id_ptr != nullptr);
+    if (*r_id_ptr == nullptr) {
+      return ID_REMAP_RESULT_SOURCE_NOT_MAPPABLE;
+    }
+
+    if (!mappings.contains(*r_id_ptr)) {
+      return ID_REMAP_RESULT_SOURCE_UNAVAILABLE;
+    }
+
+    if (options & ID_REMAP_APPLY_UPDATE_REFCOUNT) {
+      id_us_min(*r_id_ptr);
+    }
+
+    *r_id_ptr = mappings.lookup(*r_id_ptr);
+    if (*r_id_ptr == nullptr) {
+      return ID_REMAP_RESULT_SOURCE_UNASSIGNED;
+    }
+
+    if (options & ID_REMAP_APPLY_UPDATE_REFCOUNT) {
+      id_us_plus(*r_id_ptr);
+    }
+
+    if (options & ID_REMAP_APPLY_ENSURE_REAL) {
+      id_us_ensure_real(*r_id_ptr);
+    }
+    return ID_REMAP_RESULT_SOURCE_REMAPPED;
+  }
+
+  void iter(IDRemapperIterFunction func, void *user_data) const
+  {
+    for (auto item : mappings.items()) {
+      func(item.key, item.value, user_data);
+    }
+  }
+};
+
+}  // namespace blender::bke::id::remapper
+
+/** \brief wrap CPP IDRemapper to a C handle. */
+static IDRemapper *wrap(blender::bke::id::remapper::IDRemapper *remapper)
+{
+  return static_cast<IDRemapper *>(static_cast<void *>(remapper));
+}
+
+/** \brief wrap C handle to a CPP IDRemapper. */
+static blender::bke::id::remapper::IDRemapper *unwrap(IDRemapper *remapper)
+{
+  return static_cast<blender::bke::id::remapper::IDRemapper *>(static_cast<void *>(remapper));
+}
+
+/** \brief wrap C handle to a CPP IDRemapper. */
+static const blender::bke::id::remapper::IDRemapper *unwrap(const IDRemapper *remapper)
+{
+  return static_cast<const blender::bke::id::remapper::IDRemapper *>(
+      static_cast<const void *>(remapper));
+}
+
+extern "C" {
+
+IDRemapper *BKE_id_remapper_create(void)
+{
+  blender::bke::id::remapper::IDRemapper *remapper =
+      MEM_new<blender::bke::id::remapper::IDRemapper>(__func__);
+  return wrap(remapper);
+}
+
+void BKE_id_remapper_free(IDRemapper *id_remapper)
+{
+  blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  MEM_delete<blender::bke::id::remapper::IDRemapper>(remapper);
+}
+
+void BKE_id_remapper_clear(struct IDRemapper *id_remapper)
+{
+  blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  remapper->clear();
+}
+
+bool BKE_id_remapper_is_empty(const struct IDRemapper *id_remapper)
+{
+  const blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  return remapper->is_empty();
+}
+
+void BKE_id_remapper_add(IDRemapper *id_remapper, ID *old_id, ID *new_id)
+{
+  blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  remapper->add(old_id, new_id);
+}
+
+bool BKE_id_remapper_has_mapping_for(const struct IDRemapper *id_remapper, uint64_t type_filter)
+{
+  const blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  return remapper->contains_mappings_for_any(type_filter);
+}
+
+IDRemapperApplyResult BKE_id_remapper_apply(const IDRemapper *id_remapper,
+                                            ID **r_id_ptr,
+                                            const IDRemapperApplyOptions options)
+{
+  const blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  return remapper->apply(r_id_ptr, options);
+}
+
+void BKE_id_remapper_iter(const struct IDRemapper *id_remapper,
+                          IDRemapperIterFunction func,
+                          void *user_data)
+{
+  const blender::bke::id::remapper::IDRemapper *remapper = unwrap(id_remapper);
+  remapper->iter(func, user_data);
+}
+}

source/blender/blenkernel/intern/lib_id_remapper_test.cc

@@ -0,0 +1,83 @@
+/*
+ * 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) 2022 by Blender Foundation.
+ */
+
+#include "testing/testing.h"
+
+#include "BKE_lib_remap.h"
+
+#include "BLI_string.h"
+
+#include "DNA_ID.h"
+
+namespace blender::bke::id::remapper::tests {
+
+TEST(lib_id_remapper, unavailable)
+{
+  ID id1;
+  ID *idp = &id1;
+
+  IDRemapper *remapper = BKE_id_remapper_create();
+  IDRemapperApplyResult result = BKE_id_remapper_apply(remapper, &idp, ID_REMAP_APPLY_DEFAULT);
+  EXPECT_EQ(result, ID_REMAP_RESULT_SOURCE_UNAVAILABLE);
+
+  BKE_id_remapper_free(remapper);
+}
+
+TEST(lib_id_remapper, not_mappable)
+{
+  ID *idp = nullptr;
+
+  IDRemapper *remapper = BKE_id_remapper_create();
+  IDRemapperApplyResult result = BKE_id_remapper_apply(remapper, &idp, ID_REMAP_APPLY_DEFAULT);
+  EXPECT_EQ(result, ID_REMAP_RESULT_SOURCE_NOT_MAPPABLE);
+
+  BKE_id_remapper_free(remapper);
+}
+
+TEST(lib_id_remapper, mapped)
+{
+  ID id1;
+  ID id2;
+  ID *idp = &id1;
+  BLI_strncpy(id1.name, "OB1", sizeof(id1.name));
+  BLI_strncpy(id2.name, "OB2", sizeof(id2.name));
+
+  IDRemapper *remapper = BKE_id_remapper_create();
+  BKE_id_remapper_add(remapper, &id1, &id2);
+  IDRemapperApplyResult result = BKE_id_remapper_apply(remapper, &idp, ID_REMAP_APPLY_DEFAULT);
+  EXPECT_EQ(result, ID_REMAP_RESULT_SOURCE_REMAPPED);
+  EXPECT_EQ(idp, &id2);
+
+  BKE_id_remapper_free(remapper);
+}
+
+TEST(lib_id_remapper, unassigned)
+{
+  ID id1;
+  ID *idp = &id1;
+
+  IDRemapper *remapper = BKE_id_remapper_create();
+  BKE_id_remapper_add(remapper, &id1, nullptr);
+  IDRemapperApplyResult result = BKE_id_remapper_apply(remapper, &idp, ID_REMAP_APPLY_DEFAULT);
+  EXPECT_EQ(result, ID_REMAP_RESULT_SOURCE_UNASSIGNED);
+  EXPECT_EQ(idp, nullptr);
+
+  BKE_id_remapper_free(remapper);
+}
+
+}  // namespace blender::bke::id::remapper::tests

source/blender/blenkernel/intern/lib_override.c

@@ -1121,6 +1121,45 @@ void BKE_lib_override_library_main_proxy_convert(Main *bmain, BlendFileReadRepor
   }
 }
 
+static void lib_override_library_remap(Main *bmain,
+                                       const ID *id_root_reference,
+                                       GHash *linkedref_to_old_override)
+{
+  ID *id;
+  struct IDRemapper *remapper = BKE_id_remapper_create();
+  FOREACH_MAIN_ID_BEGIN (bmain, id) {
+
+    if (id->tag & LIB_TAG_DOIT && id->newid != NULL && id->lib == id_root_reference->lib) {
+      ID *id_override_new = id->newid;
+      ID *id_override_old = BLI_ghash_lookup(linkedref_to_old_override, id);
+      if (id_override_old == NULL) {
+        continue;
+      }
+
+      BKE_id_remapper_add(remapper, id_override_old, id_override_new);
+      /* Remap no-main override IDs we just created too. */
+      GHashIterator linkedref_to_old_override_iter;
+      GHASH_ITER (linkedref_to_old_override_iter, linkedref_to_old_override) {
+        ID *id_override_old_iter = BLI_ghashIterator_getValue(&linkedref_to_old_override_iter);
+        if ((id_override_old_iter->tag & LIB_TAG_NO_MAIN) == 0) {
+          continue;
+        }
+
+        BKE_libblock_relink_ex(bmain,
+                               id_override_old_iter,
+                               id_override_old,
+                               id_override_new,
+                               ID_REMAP_FORCE_USER_REFCOUNT | ID_REMAP_FORCE_NEVER_NULL_USAGE);
+      }
+    }
+  }
+  FOREACH_MAIN_ID_END;
+
+  /* Remap all IDs to use the new override. */
+  BKE_libblock_remap_multiple(bmain, remapper, 0);
+  BKE_id_remapper_free(remapper);
+}
+
 static bool lib_override_library_resync(Main *bmain,
                                         Scene *scene,
                                         ViewLayer *view_layer,
@@ -1312,32 +1351,9 @@ static bool lib_override_library_resync(Main *bmain,
   }
   FOREACH_MAIN_LISTBASE_END;
 
-  /* We need to remap old to new override usages in a separate loop, after all new overrides have
+  /* We remap old to new override usages in a separate loop, after all new overrides have
    * been added to Main. */
-  FOREACH_MAIN_ID_BEGIN (bmain, id) {
-    if (id->tag & LIB_TAG_DOIT && id->newid != NULL && id->lib == id_root_reference->lib) {
-      ID *id_override_new = id->newid;
-      ID *id_override_old = BLI_ghash_lookup(linkedref_to_old_override, id);
-
-      if (id_override_old != NULL) {
-        /* Remap all IDs to use the new override. */
-        BKE_libblock_remap(bmain, id_override_old, id_override_new, 0);
-        /* Remap no-main override IDs we just created too. */
-        GHashIterator linkedref_to_old_override_iter;
-        GHASH_ITER (linkedref_to_old_override_iter, linkedref_to_old_override) {
-          ID *id_override_old_iter = BLI_ghashIterator_getValue(&linkedref_to_old_override_iter);
-          if (id_override_old_iter->tag & LIB_TAG_NO_MAIN) {
-            BKE_libblock_relink_ex(bmain,
-                                   id_override_old_iter,
-                                   id_override_old,
-                                   id_override_new,
-                                   ID_REMAP_FORCE_USER_REFCOUNT | ID_REMAP_FORCE_NEVER_NULL_USAGE);
-          }
-        }
-      }
-    }
-  }
-  FOREACH_MAIN_ID_END;
+  lib_override_library_remap(bmain, id_root_reference, linkedref_to_old_override);
 
   BKE_main_collection_sync(bmain);
 
@@ -1537,11 +1553,13 @@ static void lib_override_resync_tagging_finalize_recurse(Main *bmain,
     CLOG_ERROR(
         &LOG,
         "While processing indirect level %d, ID %s from lib %s of indirect level %d detected "
-        "as needing resync.",
+        "as needing resync, skipping.",
         library_indirect_level,
         id->name,
         id->lib->filepath,
         id->lib->temp_index);
+    id->tag &= ~LIB_TAG_LIB_OVERRIDE_NEED_RESYNC;
+    return;
   }
 
   MainIDRelationsEntry *entry = BLI_ghash_lookup(bmain->relations->relations_from_pointers, id);

source/blender/blenkernel/intern/lib_remap.c

@@ -510,11 +510,18 @@ static void libblock_remap_data(
 #endif
 }
 
-void BKE_libblock_remap_locked(Main *bmain, void *old_idv, void *new_idv, const short remap_flags)
+typedef struct LibblockRemapMultipleUserData {
+  Main *bmain;
+  short remap_flags;
+} LibBlockRemapMultipleUserData;
+
+static void libblock_remap_foreach_idpair_cb(ID *old_id, ID *new_id, void *user_data)
 {
+  LibBlockRemapMultipleUserData *data = user_data;
+  Main *bmain = data->bmain;
+  const short remap_flags = data->remap_flags;
+
   IDRemap id_remap_data;
-  ID *old_id = old_idv;
-  ID *new_id = new_idv;
   int skipped_direct, skipped_refcounted;
 
   BLI_assert(old_id != NULL);
@@ -527,13 +534,6 @@ void BKE_libblock_remap_locked(Main *bmain, void *old_idv, void *new_idv, const
     free_notifier_reference_cb(old_id);
   }
 
-  /* We assume editors do not hold references to their IDs... This is false in some cases
-   * (Image is especially tricky here),
-   * editors' code is to handle refcount (id->us) itself then. */
-  if (remap_editor_id_reference_cb) {
-    remap_editor_id_reference_cb(old_id, new_id);
-  }
-
   skipped_direct = id_remap_data.skipped_direct;
   skipped_refcounted = id_remap_data.skipped_refcounted;
 
@@ -606,6 +606,41 @@ void BKE_libblock_remap_locked(Main *bmain, void *old_idv, void *new_idv, const
   DEG_relations_tag_update(bmain);
 }
 
+void BKE_libblock_remap_multiple_locked(Main *bmain,
+                                        const struct IDRemapper *mappings,
+                                        const short remap_flags)
+{
+  if (BKE_id_remapper_is_empty(mappings)) {
+    /* Early exit nothing to do. */
+    return;
+  }
+
+  LibBlockRemapMultipleUserData user_data;
+  user_data.bmain = bmain;
+  user_data.remap_flags = remap_flags;
+  BKE_id_remapper_iter(mappings, libblock_remap_foreach_idpair_cb, &user_data);
+
+  /* We assume editors do not hold references to their IDs... This is false in some cases
+   * (Image is especially tricky here),
+   * editors' code is to handle refcount (id->us) itself then. */
+  if (remap_editor_id_reference_cb) {
+    remap_editor_id_reference_cb(mappings);
+  }
+
+  /* Full rebuild of DEG! */
+  DEG_relations_tag_update(bmain);
+}
+
+void BKE_libblock_remap_locked(Main *bmain, void *old_idv, void *new_idv, const short remap_flags)
+{
+  struct IDRemapper *remapper = BKE_id_remapper_create();
+  ID *old_id = old_idv;
+  ID *new_id = new_idv;
+  BKE_id_remapper_add(remapper, old_id, new_id);
+  BKE_libblock_remap_multiple_locked(bmain, remapper, remap_flags);
+  BKE_id_remapper_free(remapper);
+}
+
 void BKE_libblock_remap(Main *bmain, void *old_idv, void *new_idv, const short remap_flags)
 {
   BKE_main_lock(bmain);
@@ -615,6 +650,17 @@ void BKE_libblock_remap(Main *bmain, void *old_idv, void *new_idv, const short r
   BKE_main_unlock(bmain);
 }
 
+void BKE_libblock_remap_multiple(Main *bmain,
+                                 const struct IDRemapper *mappings,
+                                 const short remap_flags)
+{
+  BKE_main_lock(bmain);
+
+  BKE_libblock_remap_multiple_locked(bmain, mappings, remap_flags);
+
+  BKE_main_unlock(bmain);
+}
+
 void BKE_libblock_unlink(Main *bmain,
                          void *idv,
                          const bool do_flag_never_null,

source/blender/blenkernel/intern/mesh_iterators.c

@@ -46,7 +46,7 @@ void BKE_mesh_foreach_mapped_vert(
     BMIter iter;
     BMVert *eve;
     int i;
-    if (mesh->runtime.edit_data->vertexCos != NULL) {
+    if (mesh->runtime.edit_data != NULL && mesh->runtime.edit_data->vertexCos != NULL) {
       const float(*vertexCos)[3] = mesh->runtime.edit_data->vertexCos;
       const float(*vertexNos)[3];
       if (flag & MESH_FOREACH_USE_NORMAL) {
@@ -106,7 +106,7 @@ void BKE_mesh_foreach_mapped_edge(
     BMIter iter;
     BMEdge *eed;
     int i;
-    if (mesh->runtime.edit_data->vertexCos != NULL) {
+    if (mesh->runtime.edit_data != NULL && mesh->runtime.edit_data->vertexCos != NULL) {
       const float(*vertexCos)[3] = mesh->runtime.edit_data->vertexCos;
       BM_mesh_elem_index_ensure(bm, BM_VERT);
 
@@ -164,7 +164,8 @@ void BKE_mesh_foreach_mapped_loop(Mesh *mesh,
     BMIter iter;
     BMFace *efa;
 
-    const float(*vertexCos)[3] = mesh->runtime.edit_data->vertexCos;
+    const float(*vertexCos)[3] = mesh->runtime.edit_data ? mesh->runtime.edit_data->vertexCos :
+                                                           NULL;
 
     /* XXX: investigate using EditMesh data. */
     const float(*lnors)[3] = (flag & MESH_FOREACH_USE_NORMAL) ?
@@ -231,7 +232,7 @@ void BKE_mesh_foreach_mapped_face_center(
     void *userData,
     MeshForeachFlag flag)
 {
-  if (mesh->edit_mesh != NULL) {
+  if (mesh->edit_mesh != NULL && mesh->runtime.edit_data != NULL) {
     BMEditMesh *em = mesh->edit_mesh;
     BMesh *bm = em->bm;
     const float(*polyCos)[3];

source/blender/blenlib/BLI_string_ref.hh

@@ -337,6 +337,18 @@ constexpr int64_t StringRefBase::find(StringRef str, int64_t pos) const
   return index_or_npos_to_int64(std::string_view(*this).find(str, static_cast<size_t>(pos)));
 }
 
+constexpr int64_t StringRefBase::rfind(char c, int64_t pos) const
+{
+  BLI_assert(pos >= 0);
+  return index_or_npos_to_int64(std::string_view(*this).rfind(c, static_cast<size_t>(pos)));
+}
+
+constexpr int64_t StringRefBase::rfind(StringRef str, int64_t pos) const
+{
+  BLI_assert(pos >= 0);
+  return index_or_npos_to_int64(std::string_view(*this).rfind(str, static_cast<size_t>(pos)));
+}
+
 constexpr int64_t StringRefBase::find_first_of(StringRef chars, int64_t pos) const
 {
   BLI_assert(pos >= 0);

source/blender/draw/CMakeLists.txt

@@ -154,7 +154,7 @@ set(SRC
   engines/workbench/workbench_materials.c
   engines/workbench/workbench_opaque.c
   engines/workbench/workbench_render.c
-  engines/workbench/workbench_shader.c
+  engines/workbench/workbench_shader.cc
   engines/workbench/workbench_shadow.c
   engines/workbench/workbench_transparent.c
   engines/workbench/workbench_volume.c
@@ -342,7 +342,6 @@ set(GLSL_SRC
   engines/workbench/shaders/workbench_common_lib.glsl
   engines/workbench/shaders/workbench_composite_frag.glsl
   engines/workbench/shaders/workbench_curvature_lib.glsl
-  engines/workbench/shaders/workbench_data_lib.glsl
   engines/workbench/shaders/workbench_effect_cavity_frag.glsl
   engines/workbench/shaders/workbench_effect_dof_frag.glsl
   engines/workbench/shaders/workbench_effect_outline_frag.glsl
@@ -357,7 +356,6 @@ set(GLSL_SRC
   engines/workbench/shaders/workbench_prepass_hair_vert.glsl
   engines/workbench/shaders/workbench_prepass_pointcloud_vert.glsl
   engines/workbench/shaders/workbench_prepass_vert.glsl
-  engines/workbench/shaders/workbench_shader_interface_lib.glsl
   engines/workbench/shaders/workbench_shadow_caps_geom.glsl
   engines/workbench/shaders/workbench_shadow_debug_frag.glsl
   engines/workbench/shaders/workbench_shadow_geom.glsl
@@ -378,6 +376,7 @@ set(GLSL_SRC
   intern/shaders/common_hair_refine_comp.glsl
   intern/shaders/common_math_lib.glsl
   intern/shaders/common_math_geom_lib.glsl
+  intern/shaders/common_view_clipping_lib.glsl
   intern/shaders/common_view_lib.glsl
   intern/shaders/common_fxaa_lib.glsl
   intern/shaders/common_smaa_lib.glsl
@@ -532,7 +531,7 @@ set(GLSL_SOURCE_CONTENT "")
 foreach(GLSL_FILE ${GLSL_SRC})
   get_filename_component(GLSL_FILE_NAME ${GLSL_FILE} NAME)
   string(REPLACE "." "_" GLSL_FILE_NAME_UNDERSCORES ${GLSL_FILE_NAME})
-  string(APPEND GLSL_SOURCE_CONTENT "SHADER_SOURCE\(datatoc_${GLSL_FILE_NAME_UNDERSCORES}, \"${GLSL_FILE_NAME}\"\)\n")
+  string(APPEND GLSL_SOURCE_CONTENT "SHADER_SOURCE\(datatoc_${GLSL_FILE_NAME_UNDERSCORES}, \"${GLSL_FILE_NAME}\", \"${GLSL_FILE}\"\)\n")
 endforeach()
 
 set(glsl_source_list_file "${CMAKE_CURRENT_BINARY_DIR}/glsl_draw_source_list.h")

source/blender/draw/engines/workbench/shaders/infos/workbench_composite_info.hh

@@ -0,0 +1,41 @@
+
+#include "gpu_shader_create_info.hh"
+
+/* -------------------------------------------------------------------- */
+/** \name Base Composite
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_composite)
+    .sampler(0, ImageType::FLOAT_2D, "normalBuffer", Frequency::PASS)
+    .sampler(1, ImageType::FLOAT_2D, "materialBuffer", Frequency::PASS)
+    .uniform_buf(4, "WorldData", "world_data", Frequency::PASS)
+    .push_constant(0, Type::BOOL, "forceShadowing")
+    .fragment_out(0, Type::VEC4, "fragColor")
+    .typedef_source("workbench_shader_shared.h")
+    .fragment_source("workbench_composite_frag.glsl")
+    .additional_info("draw_fullscreen");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Lighting Type
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_composite_studio)
+    .define("V3D_LIGHTING_STUDIO")
+    .additional_info("workbench_composite")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_composite_matcap)
+    .define("V3D_LIGHTING_MATCAP")
+    .sampler(2, ImageType::FLOAT_2D, "matcap_diffuse_tx", Frequency::PASS)
+    .sampler(3, ImageType::FLOAT_2D, "matcap_specular_tx", Frequency::PASS)
+    .additional_info("workbench_composite")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_composite_flat)
+    .define("V3D_LIGHTING_FLAT")
+    .additional_info("workbench_composite")
+    .do_static_compilation(true);
+
+/** \} */

source/blender/draw/engines/workbench/shaders/infos/workbench_effect_antialiasing_info.hh

@@ -0,0 +1,64 @@
+
+#include "gpu_shader_create_info.hh"
+
+/* -------------------------------------------------------------------- */
+/** \name TAA
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_taa)
+    .sampler(0, ImageType::FLOAT_2D, "colorBuffer")
+    .push_constant(0, Type::FLOAT, "samplesWeights", 9)
+    .fragment_out(0, Type::VEC4, "fragColor")
+    .fragment_source("workbench_effect_taa_frag.glsl")
+    .additional_info("draw_fullscreen")
+    .do_static_compilation(true);
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name SMAA
+ * \{ */
+
+GPU_SHADER_INTERFACE_INFO(workbench_smaa_iface, "")
+    .smooth(Type::VEC2, "uvs")
+    .smooth(Type::VEC2, "pixcoord")
+    .smooth(Type::VEC4, "offset[3]");
+
+GPU_SHADER_CREATE_INFO(workbench_smaa)
+    .define("SMAA_GLSL_3")
+    .define("SMAA_RT_METRICS", "viewportMetrics")
+    .define("SMAA_PRESET_HIGH")
+    .define("SMAA_LUMA_WEIGHT", "float4(1.0, 1.0, 1.0, 1.0)")
+    .define("SMAA_NO_DISCARD")
+    .vertex_out(workbench_smaa_iface)
+    .push_constant(1, Type::VEC4, "viewportMetrics")
+    .vertex_source("workbench_effect_smaa_vert.glsl")
+    .fragment_source("workbench_effect_smaa_frag.glsl");
+
+GPU_SHADER_CREATE_INFO(workbench_smaa_stage_0)
+    .define("SMAA_STAGE", "0")
+    .sampler(0, ImageType::FLOAT_2D, "colorTex")
+    .fragment_out(0, Type::VEC2, "out_edges")
+    .additional_info("workbench_smaa")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_smaa_stage_1)
+    .define("SMAA_STAGE", "1")
+    .sampler(0, ImageType::FLOAT_2D, "edgesTex")
+    .sampler(1, ImageType::FLOAT_2D, "areaTex")
+    .sampler(2, ImageType::FLOAT_2D, "searchTex")
+    .fragment_out(0, Type::VEC4, "out_weights")
+    .additional_info("workbench_smaa")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_smaa_stage_2)
+    .define("SMAA_STAGE", "2")
+    .sampler(0, ImageType::FLOAT_2D, "colorTex")
+    .sampler(1, ImageType::FLOAT_2D, "blendTex")
+    .push_constant(2, Type::FLOAT, "mixFactor")
+    .push_constant(3, Type::FLOAT, "taaAccumulatedWeight")
+    .fragment_out(0, Type::VEC4, "out_color")
+    .additional_info("workbench_smaa")
+    .do_static_compilation(true);
+
+/** \} */

source/blender/draw/engines/workbench/shaders/infos/workbench_effect_cavity_info.hh

@@ -3,10 +3,13 @@
 
 GPU_SHADER_CREATE_INFO(workbench_effect_cavity_common)
     .fragment_out(0, Type::VEC4, "fragColor")
-    .sampler(0, ImageType::FLOAT_2D, "depthBuffer")
+    .sampler(0, ImageType::DEPTH_2D, "depthBuffer")
     .sampler(1, ImageType::FLOAT_2D, "normalBuffer")
     .sampler(2, ImageType::UINT_2D, "objectIdBuffer")
+    .sampler(3, ImageType::FLOAT_2D, "cavityJitter")
     .uniform_buf(3, "vec4", "samples_coords[512]")
+    .uniform_buf(4, "WorldData", "world_data", Frequency::PASS)
+    .typedef_source("workbench_shader_shared.h")
     .fragment_source("workbench_effect_cavity_frag.glsl")
     .additional_info("draw_fullscreen")
     .additional_info("draw_view");

source/blender/draw/engines/workbench/shaders/infos/workbench_effect_dof_info.hh

@@ -0,0 +1,55 @@
+
+#include "gpu_shader_create_info.hh"
+
+GPU_SHADER_CREATE_INFO(workbench_effect_dof)
+    /* TODO(fclem): Split resources per stage. */
+    .sampler(0, ImageType::FLOAT_2D, "inputCocTex")
+    .sampler(1, ImageType::FLOAT_2D, "maxCocTilesTex")
+    .sampler(2, ImageType::FLOAT_2D, "sceneColorTex")
+    .sampler(3, ImageType::FLOAT_2D, "sceneDepthTex")
+    .sampler(4, ImageType::FLOAT_2D, "backgroundTex")
+    .sampler(5, ImageType::FLOAT_2D, "halfResColorTex")
+    .sampler(6, ImageType::FLOAT_2D, "blurTex")
+    .sampler(7, ImageType::FLOAT_2D, "noiseTex")
+    .push_constant(0, Type::VEC2, "invertedViewportSize")
+    .push_constant(1, Type::VEC2, "nearFar")
+    .push_constant(2, Type::VEC3, "dofParams")
+    .push_constant(3, Type::FLOAT, "noiseOffset")
+    .fragment_source("workbench_effect_dof_frag.glsl")
+    .additional_info("draw_fullscreen")
+    .additional_info("draw_view");
+
+GPU_SHADER_CREATE_INFO(workbench_effect_dof_prepare)
+    .define("PREPARE")
+    .fragment_out(0, Type::VEC4, "halfResColor")
+    .fragment_out(1, Type::VEC2, "normalizedCoc")
+    .additional_info("workbench_effect_dof")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_effect_dof_downsample)
+    .define("DOWNSAMPLE")
+    .fragment_out(0, Type::VEC4, "outColor")
+    .fragment_out(1, Type::VEC2, "outCocs")
+    .additional_info("workbench_effect_dof")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_effect_dof_blur1)
+    .define("BLUR1")
+    .define("NUM_SAMPLES", "49")
+    .uniform_buf(1, "vec4", "samples[49]")
+    .fragment_out(0, Type::VEC4, "blurColor")
+    .additional_info("workbench_effect_dof")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_effect_dof_blur2)
+    .define("BLUR2")
+    .fragment_out(0, Type::VEC4, "finalColor")
+    .additional_info("workbench_effect_dof")
+    .do_static_compilation(true);
+
+GPU_SHADER_CREATE_INFO(workbench_effect_dof_resolve)
+    .define("RESOLVE")
+    .fragment_out(0, Type::VEC4, "finalColorAdd", DualBlend::SRC_0)
+    .fragment_out(0, Type::VEC4, "finalColorMul", DualBlend::SRC_1)
+    .additional_info("workbench_effect_dof")
+    .do_static_compilation(true);

source/blender/draw/engines/workbench/shaders/infos/workbench_effect_outline_info.hh

@@ -0,0 +1,11 @@
+
+#include "gpu_shader_create_info.hh"
+
+GPU_SHADER_CREATE_INFO(workbench_effect_outline)
+    .typedef_source("workbench_shader_shared.h")
+    .fragment_source("workbench_effect_outline_frag.glsl")
+    .sampler(0, ImageType::UINT_2D, "objectIdBuffer")
+    .uniform_buf(4, "WorldData", "world_data", Frequency::PASS)
+    .fragment_out(0, Type::VEC4, "fragColor")
+    .additional_info("draw_fullscreen")
+    .do_static_compilation(true);

source/blender/draw/engines/workbench/shaders/infos/workbench_merge_infront_info.hh

@@ -0,0 +1,9 @@
+
+#include "gpu_shader_create_info.hh"
+
+GPU_SHADER_CREATE_INFO(workbench_merge_infront)
+    .fragment_out(0, Type::VEC4, "fragColor")
+    .sampler(0, ImageType::DEPTH_2D, "depthBuffer")
+    .fragment_source("workbench_merge_infront_frag.glsl")
+    .additional_info("draw_fullscreen")
+    .do_static_compilation(true);

source/blender/draw/engines/workbench/shaders/infos/workbench_prepass_info.hh

@@ -11,17 +11,20 @@ GPU_SHADER_CREATE_INFO(workbench_mesh)
     .vertex_in(2, Type::VEC4, "ac")
     .vertex_in(3, Type::VEC2, "au")
     .vertex_source("workbench_prepass_vert.glsl")
-    .additional_info("draw_mesh");
+    .additional_info("draw_mesh")
+    .additional_info("draw_resource_handle");
 
 GPU_SHADER_CREATE_INFO(workbench_hair)
     .sampler(0, ImageType::FLOAT_BUFFER, "ac", Frequency::BATCH)
     .sampler(1, ImageType::FLOAT_BUFFER, "au", Frequency::BATCH)
     .vertex_source("workbench_prepass_hair_vert.glsl")
-    .additional_info("draw_hair");
+    .additional_info("draw_hair")
+    .additional_info("draw_resource_handle");
 
 GPU_SHADER_CREATE_INFO(workbench_pointcloud)
     .vertex_source("workbench_prepass_pointcloud_vert.glsl")
-    .additional_info("draw_pointcloud");
+    .additional_info("draw_pointcloud")
+    .additional_info("draw_resource_handle");
 
 /** \} */
 
@@ -42,17 +45,21 @@ GPU_SHADER_CREATE_INFO(workbench_texture_tile)
     .sampler(3, ImageType::FLOAT_1D_ARRAY, "imageTileData", Frequency::BATCH)
     .push_constant(1, Type::BOOL, "imagePremult")
     .push_constant(2, Type::FLOAT, "imageTransparencyCutoff")
+    .define("V3D_SHADING_TEXTURE_COLOR")
     .define("TEXTURE_IMAGE_ARRAY");
 
 /** \} */
 
 /* -------------------------------------------------------------------- */
-/** \name Lighting Type
+/** \name Lighting Type (only for transparent)
  * \{ */
 
-GPU_SHADER_CREATE_INFO(workbench_lighting_studio).define("V3D_LIGHTING_STUDIO");
-GPU_SHADER_CREATE_INFO(workbench_lighting_matcap).define("V3D_LIGHTING_MATCAP");
 GPU_SHADER_CREATE_INFO(workbench_lighting_flat).define("V3D_LIGHTING_FLAT");
+GPU_SHADER_CREATE_INFO(workbench_lighting_studio).define("V3D_LIGHTING_STUDIO");
+GPU_SHADER_CREATE_INFO(workbench_lighting_matcap)
+    .define("V3D_LIGHTING_MATCAP")
+    .sampler(4, ImageType::FLOAT_2D, "matcap_diffuse_tx", Frequency::PASS)
+    .sampler(5, ImageType::FLOAT_2D, "matcap_specular_tx", Frequency::PASS);
 
 /** \} */
 
@@ -69,7 +76,12 @@ GPU_SHADER_INTERFACE_INFO(workbench_material_iface, "")
     .flat(Type::FLOAT, "roughness")
     .flat(Type::FLOAT, "metallic");
 
-GPU_SHADER_CREATE_INFO(workbench_material).vertex_out(workbench_material_iface);
+GPU_SHADER_CREATE_INFO(workbench_material)
+    .uniform_buf(4, "WorldData", "world_data", Frequency::PASS)
+    .uniform_buf(5, "vec4", "materials_data[4096]", Frequency::PASS)
+    .push_constant(4, Type::INT, "materialIndex")
+    .push_constant(5, Type::BOOL, "useMatcap")
+    .vertex_out(workbench_material_iface);
 
 /** \} */
 
@@ -83,19 +95,16 @@ GPU_SHADER_CREATE_INFO(workbench_transparent_accum)
     .fragment_out(0, Type::VEC4, "transparentAccum")
     .fragment_out(1, Type::VEC4, "revealageAccum")
     .fragment_out(2, Type::UINT, "objectId")
-    .uniform_buf(4, "WorldData", "world_data", Frequency::PASS)
+    .push_constant(3, Type::BOOL, "forceShadowing")
     .typedef_source("workbench_shader_shared.h")
-    .fragment_source("workbench_transparent_accum_frag.glsl")
-    .additional_info("workbench_material");
+    .fragment_source("workbench_transparent_accum_frag.glsl");
 
 GPU_SHADER_CREATE_INFO(workbench_opaque)
     .fragment_out(0, Type::VEC4, "materialData")
     .fragment_out(1, Type::VEC2, "normalData")
     .fragment_out(2, Type::UINT, "objectId")
-    .uniform_buf(4, "WorldData", "world_data", Frequency::PASS)
     .typedef_source("workbench_shader_shared.h")
-    .fragment_source("workbench_prepass_frag.glsl")
-    .additional_info("workbench_material");
+    .fragment_source("workbench_prepass_frag.glsl");
 
 /** \} */
 
@@ -103,34 +112,38 @@ GPU_SHADER_CREATE_INFO(workbench_opaque)
 /** \name Variations Declaration
  * \{ */
 
-#define WORKBENCH_SURFACETYPE_VARIATIONS(prefix, ...) \
-  GPU_SHADER_CREATE_INFO(prefix##_mesh) \
-      .additional_info("workbench_mesh", __VA_ARGS__) \
-      .do_static_compilation(true); \
-  GPU_SHADER_CREATE_INFO(prefix##_hair) \
-      .additional_info("workbench_hair", __VA_ARGS__) \
-      .do_static_compilation(true); \
-  GPU_SHADER_CREATE_INFO(prefix##_ptcloud) \
-      .additional_info("workbench_pointcloud", __VA_ARGS__) \
-      .do_static_compilation(true);
+#define WORKBENCH_FINAL_VARIATION(name, ...) \
+  GPU_SHADER_CREATE_INFO(name).additional_info(__VA_ARGS__).do_static_compilation(true);
+
+#define WORKBENCH_CLIPPING_VARIATIONS(prefix, ...) \
+  WORKBENCH_FINAL_VARIATION(prefix##_clip, "drw_clipped", __VA_ARGS__) \
+  WORKBENCH_FINAL_VARIATION(prefix##_no_clip, __VA_ARGS__)
+
+#define WORKBENCH_TEXTURE_VARIATIONS(prefix, ...) \
+  WORKBENCH_CLIPPING_VARIATIONS(prefix##_tex_none, "workbench_texture_none", __VA_ARGS__) \
+  WORKBENCH_CLIPPING_VARIATIONS(prefix##_tex_single, "workbench_texture_single", __VA_ARGS__) \
+  WORKBENCH_CLIPPING_VARIATIONS(prefix##_tex_tile, "workbench_texture_tile", __VA_ARGS__)
+
+#define WORKBENCH_DATATYPE_VARIATIONS(prefix, ...) \
+  WORKBENCH_TEXTURE_VARIATIONS(prefix##_mesh, "workbench_mesh", __VA_ARGS__) \
+  WORKBENCH_TEXTURE_VARIATIONS(prefix##_hair, "workbench_hair", __VA_ARGS__) \
+  WORKBENCH_TEXTURE_VARIATIONS(prefix##_ptcloud, "workbench_pointcloud", __VA_ARGS__)
 
 #define WORKBENCH_PIPELINE_VARIATIONS(prefix, ...) \
-  WORKBENCH_SURFACETYPE_VARIATIONS(prefix##_transp_studio, \
-                                   "workbench_transparent_accum", \
-                                   "workbench_lighting_studio", \
-                                   __VA_ARGS__) \
-  WORKBENCH_SURFACETYPE_VARIATIONS(prefix##_transp_matcap, \
-                                   "workbench_transparent_accum", \
-                                   "workbench_lighting_matcap", \
-                                   __VA_ARGS__) \
-  WORKBENCH_SURFACETYPE_VARIATIONS(prefix##_transp_flat, \
-                                   "workbench_transparent_accum", \
-                                   "workbench_lighting_flat", \
-                                   __VA_ARGS__) \
-  WORKBENCH_SURFACETYPE_VARIATIONS(prefix##_opaque, "workbench_opaque", __VA_ARGS__)
+  WORKBENCH_DATATYPE_VARIATIONS(prefix##_transp_studio, \
+                                "workbench_transparent_accum", \
+                                "workbench_lighting_studio", \
+                                __VA_ARGS__) \
+  WORKBENCH_DATATYPE_VARIATIONS(prefix##_transp_matcap, \
+                                "workbench_transparent_accum", \
+                                "workbench_lighting_matcap", \
+                                __VA_ARGS__) \
+  WORKBENCH_DATATYPE_VARIATIONS(prefix##_transp_flat, \
+                                "workbench_transparent_accum", \
+                                "workbench_lighting_flat", \
+                                __VA_ARGS__) \
+  WORKBENCH_DATATYPE_VARIATIONS(prefix##_opaque, "workbench_opaque", __VA_ARGS__)
 
-WORKBENCH_PIPELINE_VARIATIONS(workbench_tex_none, "workbench_texture_none")
-WORKBENCH_PIPELINE_VARIATIONS(workbench_tex_single, "workbench_texture_single")
-WORKBENCH_PIPELINE_VARIATIONS(workbench_tex_tile, "workbench_texture_tile")
+WORKBENCH_PIPELINE_VARIATIONS(workbench, "workbench_material");
 
 /** \} */

source/blender/draw/engines/workbench/shaders/infos/workbench_shadow_info.hh

@@ -0,0 +1,98 @@
+
+#include "gpu_shader_create_info.hh"
+
+/* -------------------------------------------------------------------- */
+/** \name Common
+ * \{ */
+
+GPU_SHADER_INTERFACE_INFO(workbench_shadow_iface, "vData")
+    .smooth(Type::VEC3, "pos")
+    .smooth(Type::VEC4, "frontPosition")
+    .smooth(Type::VEC4, "backPosition");
+
+GPU_SHADER_CREATE_INFO(workbench_shadow_common)
+    .vertex_in(0, Type::VEC3, "pos")
+    .vertex_out(workbench_shadow_iface)
+    .push_constant(0, Type::FLOAT, "lightDistance")
+    .push_constant(1, Type::VEC3, "lightDirection")
+    .vertex_source("workbench_shadow_vert.glsl")
+    .additional_info("draw_mesh");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Manifold Type
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_shadow_manifold)
+    .geometry_layout(PrimitiveIn::LINES_ADJACENCY, PrimitiveOut::TRIANGLE_STRIP, 4, 1)
+    .geometry_source("workbench_shadow_geom.glsl");
+
+GPU_SHADER_CREATE_INFO(workbench_shadow_no_manifold)
+    .geometry_layout(PrimitiveIn::LINES_ADJACENCY, PrimitiveOut::TRIANGLE_STRIP, 4, 2)
+    .geometry_source("workbench_shadow_geom.glsl");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Caps Type
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_shadow_caps)
+    .geometry_layout(PrimitiveIn::TRIANGLES, PrimitiveOut::TRIANGLE_STRIP, 3, 2)
+    .geometry_source("workbench_shadow_caps_geom.glsl");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Debug Type
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_shadow_no_debug)
+    .fragment_source("gpu_shader_depth_only_frag.glsl");
+
+GPU_SHADER_CREATE_INFO(workbench_shadow_debug)
+    .fragment_out(0, Type::VEC4, "materialData")
+    .fragment_out(1, Type::VEC4, "normalData")
+    .fragment_out(2, Type::UINT, "objectId")
+    .fragment_source("workbench_shadow_debug_frag.glsl");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Variations Declaration
+ * \{ */
+
+#define WORKBENCH_SHADOW_VARIATIONS(suffix, ...) \
+  GPU_SHADER_CREATE_INFO(workbench_shadow_pass_manifold_no_caps##suffix) \
+      .define("SHADOW_PASS") \
+      .additional_info("workbench_shadow_common", "workbench_shadow_manifold", __VA_ARGS__) \
+      .do_static_compilation(true); \
+  GPU_SHADER_CREATE_INFO(workbench_shadow_pass_no_manifold_no_caps##suffix) \
+      .define("SHADOW_PASS") \
+      .define("DOUBLE_MANIFOLD") \
+      .additional_info("workbench_shadow_common", "workbench_shadow_no_manifold", __VA_ARGS__) \
+      .do_static_compilation(true); \
+  GPU_SHADER_CREATE_INFO(workbench_shadow_fail_manifold_caps##suffix) \
+      .define("SHADOW_FAIL") \
+      .additional_info("workbench_shadow_common", "workbench_shadow_caps", __VA_ARGS__) \
+      .do_static_compilation(true); \
+  GPU_SHADER_CREATE_INFO(workbench_shadow_fail_manifold_no_caps##suffix) \
+      .define("SHADOW_FAIL") \
+      .additional_info("workbench_shadow_common", "workbench_shadow_manifold", __VA_ARGS__) \
+      .do_static_compilation(true); \
+  GPU_SHADER_CREATE_INFO(workbench_shadow_fail_no_manifold_caps##suffix) \
+      .define("SHADOW_FAIL") \
+      .define("DOUBLE_MANIFOLD") \
+      .additional_info("workbench_shadow_common", "workbench_shadow_caps", __VA_ARGS__) \
+      .do_static_compilation(true); \
+  GPU_SHADER_CREATE_INFO(workbench_shadow_fail_no_manifold_no_caps##suffix) \
+      .define("SHADOW_FAIL") \
+      .define("DOUBLE_MANIFOLD") \
+      .additional_info("workbench_shadow_common", "workbench_shadow_no_manifold", __VA_ARGS__) \
+      .do_static_compilation(true);
+
+WORKBENCH_SHADOW_VARIATIONS(, "workbench_shadow_no_debug")
+WORKBENCH_SHADOW_VARIATIONS(_debug, "workbench_shadow_debug")
+
+/** \} */

source/blender/draw/engines/workbench/shaders/infos/workbench_transparent_resolve_info.hh

@@ -0,0 +1,10 @@
+
+#include "gpu_shader_create_info.hh"
+
+GPU_SHADER_CREATE_INFO(workbench_transparent_resolve)
+    .fragment_out(0, Type::VEC4, "fragColor")
+    .sampler(0, ImageType::FLOAT_2D, "transparentAccum")
+    .sampler(1, ImageType::FLOAT_2D, "transparentRevealage")
+    .fragment_source("workbench_transparent_resolve_frag.glsl")
+    .additional_info("draw_fullscreen")
+    .do_static_compilation(true);

source/blender/draw/engines/workbench/shaders/infos/workbench_volume_info.hh

@@ -0,0 +1,114 @@
+
+#include "gpu_shader_create_info.hh"
+
+/* -------------------------------------------------------------------- */
+/** \name Volume shader base
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_volume)
+    .vertex_in(0, Type::VEC3, "pos")
+    .fragment_out(0, Type::VEC4, "fragColor")
+    .sampler(0, ImageType::DEPTH_2D, "depthBuffer")
+    .sampler(1, ImageType::FLOAT_3D, "densityTexture")
+    .push_constant(28, Type::INT, "samplesLen")
+    .push_constant(29, Type::FLOAT, "noiseOfs")
+    .push_constant(30, Type::FLOAT, "stepLength")
+    .push_constant(31, Type::FLOAT, "densityScale")
+    .vertex_source("workbench_volume_vert.glsl")
+    .fragment_source("workbench_volume_frag.glsl")
+    .additional_info("draw_object_infos");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Smoke variation
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_volume_smoke)
+    .define("VOLUME_SMOKE")
+    .sampler(2, ImageType::FLOAT_3D, "flameTexture")
+    .sampler(3, ImageType::FLOAT_1D, "flameColorTexture")
+    .additional_info("draw_mesh", "draw_resource_id_varying");
+
+GPU_SHADER_CREATE_INFO(workbench_volume_object)
+    .define("VOLUME_OBJECT")
+    .push_constant(0, Type::MAT4, "volumeTextureToObject")
+    /* FIXME(fclem): This overflow the push_constant limit. */
+    .push_constant(16, Type::MAT4, "volumeObjectToTexture")
+    .additional_info("draw_volume", "draw_resource_id_varying");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Color Band variation
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_volume_coba)
+    .define("USE_COBA")
+    .sampler(4, ImageType::UINT_3D, "flagTexture")
+    .sampler(5, ImageType::FLOAT_1D, "transferTexture")
+    .push_constant(18, Type::BOOL, "showPhi")
+    .push_constant(19, Type::BOOL, "showFlags")
+    .push_constant(20, Type::BOOL, "showPressure")
+    .push_constant(21, Type::FLOAT, "gridScale");
+
+GPU_SHADER_CREATE_INFO(workbench_volume_no_coba)
+    .sampler(4, ImageType::FLOAT_3D, "shadowTexture")
+    .sampler(5, ImageType::UINT_2D, "transferTexture")
+    .push_constant(18, Type::VEC3, "activeColor");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Sampling variation
+ * \{ */
+
+GPU_SHADER_CREATE_INFO(workbench_volume_linear).define("USE_TRILINEAR");
+GPU_SHADER_CREATE_INFO(workbench_volume_cubic).define("USE_TRICUBIC");
+GPU_SHADER_CREATE_INFO(workbench_volume_closest).define("USE_CLOSEST");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Slice variation
+ * \{ */
+
+GPU_SHADER_INTERFACE_INFO(workbench_volume_iface, "").smooth(Type::VEC3, "localPos");
+
+GPU_SHADER_CREATE_INFO(workbench_volume_slice)
+    .define("VOLUME_SLICE")
+    .vertex_in(1, Type::VEC3, "uvs")
+    .vertex_out(workbench_volume_iface)
+    .push_constant(32, Type::INT, "sliceAxis") /* -1 is no slice. */
+    .push_constant(33, Type::FLOAT, "slicePosition");
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Variations Declaration
+ * \{ */
+
+#define WORKBENCH_VOLUME_SLICE_VARIATIONS(prefix, ...) \
+  GPU_SHADER_CREATE_INFO(prefix##_slice) \
+      .additional_info("workbench_volume_slice", __VA_ARGS__) \
+      .do_static_compilation(true); \
+  GPU_SHADER_CREATE_INFO(prefix##_no_slice) \
+      .additional_info(__VA_ARGS__) \
+      .do_static_compilation(true);
+
+#define WORKBENCH_VOLUME_COBA_VARIATIONS(prefix, ...) \
+  WORKBENCH_VOLUME_SLICE_VARIATIONS(prefix##_coba, "workbench_volume_coba", __VA_ARGS__) \
+  WORKBENCH_VOLUME_SLICE_VARIATIONS(prefix##_no_coba, "workbench_volume_no_coba", __VA_ARGS__)
+
+#define WORKBENCH_VOLUME_INTERP_VARIATIONS(prefix, ...) \
+  WORKBENCH_VOLUME_COBA_VARIATIONS(prefix##_linear, "workbench_volume_linear", __VA_ARGS__) \
+  WORKBENCH_VOLUME_COBA_VARIATIONS(prefix##_cubic, "workbench_volume_cubic", __VA_ARGS__) \
+  WORKBENCH_VOLUME_COBA_VARIATIONS(prefix##_closest, "workbench_volume_closest", __VA_ARGS__)
+
+#define WORKBENCH_VOLUME_SMOKE_VARIATIONS(prefix, ...) \
+  WORKBENCH_VOLUME_INTERP_VARIATIONS(prefix##_smoke, "workbench_volume_smoke", __VA_ARGS__) \
+  WORKBENCH_VOLUME_INTERP_VARIATIONS(prefix##_object, "workbench_volume_object", __VA_ARGS__)
+
+WORKBENCH_VOLUME_SMOKE_VARIATIONS(workbench_volume, "workbench_volume")
+
+/** \} */

source/blender/draw/engines/workbench/shaders/workbench_cavity_lib.glsl

@@ -1,15 +1,7 @@
 
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_data_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 
-layout(std140) uniform samples_block
-{
-  vec4 samples_coords[512];
-};
-
-uniform sampler2D cavityJitter;
-
 /*  From The Alchemy screen-space ambient obscurance algorithm
  * http://graphics.cs.williams.edu/papers/AlchemyHPG11/VV11AlchemyAO.pdf */
 

source/blender/draw/engines/workbench/shaders/workbench_composite_frag.glsl

@@ -4,13 +4,6 @@
 #pragma BLENDER_REQUIRE(workbench_matcap_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_world_light_lib.glsl)
 
-uniform sampler2D materialBuffer;
-uniform sampler2D normalBuffer;
-
-in vec4 uvcoordsvar;
-
-out vec4 fragColor;
-
 void main()
 {
   /* Normal and Incident vector are in viewspace. Lighting is evaluated in viewspace. */
@@ -27,7 +20,7 @@ void main()
   /* When using matcaps, mat_data.a is the back-face sign. */
   N = (mat_data.a > 0.0) ? N : -N;
 
-  fragColor.rgb = get_matcap_lighting(base_color, N, I);
+  fragColor.rgb = get_matcap_lighting(matcap_diffuse_tx, matcap_specular_tx, base_color, N, I);
 #endif
 
 #ifdef V3D_LIGHTING_STUDIO
@@ -38,7 +31,7 @@ void main()
   fragColor.rgb = base_color;
 #endif
 
-  fragColor.rgb *= get_shadow(N);
+  fragColor.rgb *= get_shadow(N, forceShadowing);
 
   fragColor.a = 1.0;
 }

source/blender/draw/engines/workbench/shaders/workbench_curvature_lib.glsl

@@ -1,6 +1,4 @@
 
-#pragma BLENDER_REQUIRE(workbench_data_lib.glsl)
-
 float curvature_soft_clamp(float curvature, float control)
 {
   if (curvature < 0.5 / control) {

source/blender/draw/engines/workbench/shaders/workbench_data_lib.glsl

@@ -1,48 +0,0 @@
-
-#ifndef WORKBENCH_SHADER_SHARED_H
-struct LightData {
-  vec4 direction;
-  vec4 specular_color;
-  vec4 diffuse_color_wrap; /* rgb: diffuse col a: wrapped lighting factor */
-};
-
-struct WorldData {
-  vec4 viewport_size;
-  vec4 object_outline_color;
-  vec4 shadow_direction_vs;
-  float shadow_focus;
-  float shadow_shift;
-  float shadow_mul;
-  float shadow_add;
-  /* - 16 bytes alignment - */
-  LightData lights[4];
-  vec4 ambient_color;
-
-  int cavity_sample_start;
-  int cavity_sample_end;
-  float cavity_sample_count_inv;
-  float cavity_jitter_scale;
-
-  float cavity_valley_factor;
-  float cavity_ridge_factor;
-  float cavity_attenuation;
-  float cavity_distance;
-
-  float curvature_ridge;
-  float curvature_valley;
-  float ui_scale;
-  float _pad0;
-
-  int matcap_orientation;
-  bool use_specular;
-  int _pad1;
-  int _pad2;
-};
-
-#  define viewport_size_inv viewport_size.zw
-
-layout(std140) uniform world_block
-{
-  WorldData world_data;
-};
-#endif

source/blender/draw/engines/workbench/shaders/workbench_effect_cavity_frag.glsl

@@ -4,18 +4,6 @@
 #pragma BLENDER_REQUIRE(workbench_cavity_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_curvature_lib.glsl)
 
-#ifndef DRW_SHADER_SHARED_H
-
-uniform sampler2D depthBuffer;
-uniform sampler2D normalBuffer;
-uniform usampler2D objectIdBuffer;
-
-in vec4 uvcoordsvar;
-
-out vec4 fragColor;
-
-#endif
-
 void main()
 {
   float cavity = 0.0, edges = 0.0, curvature = 0.0;

source/blender/draw/engines/workbench/shaders/workbench_effect_dof_frag.glsl

@@ -7,19 +7,6 @@
  * Converted and adapted from HLSL to GLSL by Clément Foucault
  */
 
-uniform vec2 invertedViewportSize;
-uniform vec2 nearFar;
-uniform vec3 dofParams;
-uniform float noiseOffset;
-uniform sampler2D inputCocTex;
-uniform sampler2D maxCocTilesTex;
-uniform sampler2D sceneColorTex;
-uniform sampler2D sceneDepthTex;
-uniform sampler2D backgroundTex;
-uniform sampler2D halfResColorTex;
-uniform sampler2D blurTex;
-uniform sampler2D noiseTex;
-
 #define dof_aperturesize dofParams.x
 #define dof_distance dofParams.y
 #define dof_invsensorsize dofParams.z
@@ -53,9 +40,6 @@ float decode_signed_coc(vec2 cocs)
  */
 #ifdef PREPARE
 
-layout(location = 0) out vec4 halfResColor;
-layout(location = 1) out vec2 normalizedCoc;
-
 void main()
 {
   ivec4 texel = ivec4(gl_FragCoord.xyxy) * 2 + ivec4(0, 0, 1, 1);
@@ -99,9 +83,6 @@ void main()
  */
 #ifdef DOWNSAMPLE
 
-layout(location = 0) out vec4 outColor;
-layout(location = 1) out vec2 outCocs;
-
 void main()
 {
   vec4 texel = vec4(gl_FragCoord.xyxy) * 2.0 + vec4(0.0, 0.0, 1.0, 1.0);
@@ -216,14 +197,6 @@ void main()
  * Outputs vertical blur and combined blur in MRT
  */
 #ifdef BLUR1
-layout(location = 0) out vec4 blurColor;
-
-#  define NUM_SAMPLES 49
-
-layout(std140) uniform dofSamplesBlock
-{
-  vec4 samples[NUM_SAMPLES];
-};
 
 vec2 get_random_vector(float offset)
 {
@@ -308,7 +281,6 @@ void main()
  * OF THIS SOFTWARE, EVEN IF ADVISED OF THE POSSIBILITY OF SUCH DAMAGE.
  */
 #ifdef BLUR2
-out vec4 finalColor;
 
 void main()
 {
@@ -385,9 +357,6 @@ void main()
  */
 #ifdef RESOLVE
 
-layout(location = 0, index = 0) out vec4 finalColorAdd;
-layout(location = 0, index = 1) out vec4 finalColorMul;
-
 void main()
 {
   /* Fullscreen pass */

source/blender/draw/engines/workbench/shaders/workbench_effect_outline_frag.glsl

@@ -1,12 +1,4 @@
 
-#pragma BLENDER_REQUIRE(workbench_data_lib.glsl)
-
-uniform usampler2D objectIdBuffer;
-
-in vec4 uvcoordsvar;
-
-out vec4 fragColor;
-
 void main()
 {
   vec3 offset = vec3(world_data.viewport_size_inv, 0.0) * world_data.ui_scale;

source/blender/draw/engines/workbench/shaders/workbench_effect_smaa_frag.glsl

@@ -1,50 +1,35 @@
 
-uniform sampler2D edgesTex;
-uniform sampler2D areaTex;
-uniform sampler2D searchTex;
-uniform sampler2D blendTex;
-uniform sampler2D colorTex;
-uniform float mixFactor;
-uniform float taaAccumulatedWeight;
-
-in vec2 uvs;
-in vec2 pixcoord;
-in vec4 offset[3];
-
-#if SMAA_STAGE == 0
-out vec2 fragColor;
-#else
-out vec4 fragColor;
-#endif
+#pragma BLENDER_REQUIRE(common_smaa_lib.glsl)
 
 void main()
 {
 #if SMAA_STAGE == 0
   /* Detect edges in color and revealage buffer. */
-  fragColor = SMAALumaEdgeDetectionPS(uvs, offset, colorTex);
+  out_edges = SMAALumaEdgeDetectionPS(uvs, offset, colorTex);
   /* Discard if there is no edge. */
-  if (dot(fragColor, float2(1.0, 1.0)) == 0.0) {
+  if (dot(out_edges, float2(1.0, 1.0)) == 0.0) {
     discard;
   }
 
 #elif SMAA_STAGE == 1
-  fragColor = SMAABlendingWeightCalculationPS(
+  out_weights = SMAABlendingWeightCalculationPS(
       uvs, pixcoord, offset, edgesTex, areaTex, searchTex, vec4(0));
 
 #elif SMAA_STAGE == 2
-  fragColor = vec4(0.0);
+  out_color = vec4(0.0);
   if (mixFactor > 0.0) {
-    fragColor += SMAANeighborhoodBlendingPS(uvs, offset[0], colorTex, blendTex) * mixFactor;
+    out_color += SMAANeighborhoodBlendingPS(uvs, offset[0], colorTex, blendTex) * mixFactor;
   }
   if (mixFactor < 1.0) {
-    fragColor += texture(colorTex, uvs) * (1.0 - mixFactor);
+    out_color += texture(colorTex, uvs) * (1.0 - mixFactor);
   }
-  fragColor /= taaAccumulatedWeight;
-  fragColor = exp2(fragColor) - 0.5;
+  out_color /= taaAccumulatedWeight;
+  /* Exit log2 space used for Antialiasing. */
+  out_color = exp2(out_color) - 0.5;
 
   /* Avoid float precision issue. */
-  if (fragColor.a > 0.999) {
-    fragColor.a = 1.0;
+  if (out_color.a > 0.999) {
+    out_color.a = 1.0;
   }
 #endif
 }

source/blender/draw/engines/workbench/shaders/workbench_effect_smaa_vert.glsl

@@ -1,7 +1,5 @@
 
-out vec2 uvs;
-out vec2 pixcoord;
-out vec4 offset[3];
+#pragma BLENDER_REQUIRE(common_smaa_lib.glsl)
 
 void main()
 {

source/blender/draw/engines/workbench/shaders/workbench_effect_taa_frag.glsl

@@ -1,9 +1,4 @@
 
-uniform sampler2D colorBuffer;
-uniform float samplesWeights[9];
-
-out vec4 fragColor;
-
 void main()
 {
   vec2 texel_size = 1.0 / vec2(textureSize(colorBuffer, 0));

source/blender/draw/engines/workbench/shaders/workbench_image_lib.glsl

@@ -25,15 +25,6 @@ bool node_tex_tile_lookup(inout vec3 co, sampler2DArray ima, sampler1DArray map)
   return true;
 }
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-uniform sampler2DArray imageTileArray;
-uniform sampler1DArray imageTileData;
-uniform sampler2D imageTexture;
-
-uniform float imageTransparencyCutoff = 0.1;
-uniform bool imagePremult;
-#endif
-
 vec3 workbench_image_color(vec2 uvs)
 {
 #ifdef V3D_SHADING_TEXTURE_COLOR

source/blender/draw/engines/workbench/shaders/workbench_matcap_lib.glsl

@@ -1,6 +1,4 @@
 
-#pragma BLENDER_REQUIRE(workbench_data_lib.glsl)
-
 vec2 matcap_uv_compute(vec3 I, vec3 N, bool flipped)
 {
   /* Quick creation of an orthonormal basis */
@@ -15,16 +13,14 @@ vec2 matcap_uv_compute(vec3 I, vec3 N, bool flipped)
   return matcap_uv * 0.496 + 0.5;
 }
 
-uniform sampler2D matcapDiffuseImage;
-uniform sampler2D matcapSpecularImage;
-
-vec3 get_matcap_lighting(vec3 base_color, vec3 N, vec3 I)
+vec3 get_matcap_lighting(
+    sampler2D diffuse_matcap, sampler2D specular_matcap, vec3 base_color, vec3 N, vec3 I)
 {
   bool flipped = world_data.matcap_orientation != 0;
   vec2 uv = matcap_uv_compute(I, N, flipped);
 
-  vec3 diffuse = textureLod(matcapDiffuseImage, uv, 0.0).rgb;
-  vec3 specular = textureLod(matcapSpecularImage, uv, 0.0).rgb;
+  vec3 diffuse = textureLod(diffuse_matcap, uv, 0.0).rgb;
+  vec3 specular = textureLod(specular_matcap, uv, 0.0).rgb;
 
   return diffuse * base_color + specular * float(world_data.use_specular);
 }

source/blender/draw/engines/workbench/shaders/workbench_material_lib.glsl

@@ -1,17 +1,9 @@
 
-layout(std140) uniform material_block
-{
-  vec4 mat_data[4096];
-};
-
-/* If set to -1, the resource handle is used instead. */
-uniform int materialIndex;
-
 void workbench_material_data_get(
     int handle, out vec3 color, out float alpha, out float roughness, out float metallic)
 {
   handle = (materialIndex != -1) ? materialIndex : handle;
-  vec4 data = mat_data[uint(handle) & 0xFFFu];
+  vec4 data = materials_data[uint(handle) & 0xFFFu];
   color = data.rgb;
 
   uint encoded_data = floatBitsToUint(data.w);

source/blender/draw/engines/workbench/shaders/workbench_merge_infront_frag.glsl

@@ -1,10 +1,4 @@
 
-uniform sampler2D depthBuffer;
-
-in vec4 uvcoordsvar;
-
-out vec4 fragColor;
-
 void main()
 {
   float depth = texture(depthBuffer, uvcoordsvar.st).r;

source/blender/draw/engines/workbench/shaders/workbench_prepass_frag.glsl

@@ -1,22 +1,13 @@
 
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_shader_interface_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_image_lib.glsl)
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-layout(location = 0) out vec4 materialData;
-layout(location = 1) out vec2 normalData;
-layout(location = 2) out uint objectId;
-#endif
-
-uniform bool useMatcap = false;
-
 void main()
 {
   normalData = workbench_normal_encode(gl_FrontFacing, normal_interp);
 
-  materialData = vec4(color_interp, packed_rough_metal);
+  materialData = vec4(color_interp, workbench_float_pair_encode(roughness, metallic));
 
   objectId = uint(object_id);
 

source/blender/draw/engines/workbench/shaders/workbench_prepass_hair_vert.glsl

@@ -1,15 +1,10 @@
 #pragma BLENDER_REQUIRE(common_hair_lib.glsl)
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_shader_interface_lib.glsl)
+#pragma BLENDER_REQUIRE(common_view_clipping_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_material_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_image_lib.glsl)
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-uniform samplerBuffer ac; /* active color layer */
-uniform samplerBuffer au; /* active texture layer */
-#endif
-
 /* From http://libnoise.sourceforge.net/noisegen/index.html */
 float integer_noise(int n)
 {
@@ -65,19 +60,12 @@ void main()
   float hair_rand = integer_noise(hair_get_strand_id());
   vec3 nor = workbench_hair_random_normal(tan, binor, hair_rand);
 
-#ifdef USE_WORLD_CLIP_PLANES
-  world_clip_planes_calc_clip_distance(world_pos);
-#endif
+  view_clipping_distances(world_pos);
 
   uv_interp = hair_get_customdata_vec2(au);
 
   normal_interp = normalize(normal_world_to_view(nor));
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-#  ifdef OPAQUE_MATERIAL
-  float metallic, roughness;
-#  endif
-#endif
   workbench_material_data_get(resource_handle, color_interp, alpha_interp, roughness, metallic);
 
   if (materialIndex == 0) {
@@ -90,9 +78,5 @@ void main()
 
   workbench_hair_random_material(hair_rand, color_interp, roughness, metallic);
 
-#ifdef OPAQUE_MATERIAL
-  packed_rough_metal = workbench_float_pair_encode(roughness, metallic);
-#endif
-
   object_id = int(uint(resource_handle) & 0xFFFFu) + 1;
 }

source/blender/draw/engines/workbench/shaders/workbench_prepass_pointcloud_vert.glsl

@@ -1,7 +1,7 @@
 
+#pragma BLENDER_REQUIRE(common_view_clipping_lib.glsl)
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
 #pragma BLENDER_REQUIRE(common_pointcloud_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_shader_interface_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_material_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_image_lib.glsl)
@@ -15,26 +15,15 @@ void main()
 
   gl_Position = point_world_to_ndc(world_pos);
 
-#ifdef USE_WORLD_CLIP_PLANES
-  world_clip_planes_calc_clip_distance(world_pos);
-#endif
+  view_clipping_distances(world_pos);
 
   uv_interp = vec2(0.0);
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-#  ifdef OPAQUE_MATERIAL
-  float metallic, roughness;
-#  endif
-#endif
   workbench_material_data_get(resource_handle, color_interp, alpha_interp, roughness, metallic);
 
   if (materialIndex == 0) {
     color_interp = vec3(1.0);
   }
 
-#ifdef OPAQUE_MATERIAL
-  packed_rough_metal = workbench_float_pair_encode(roughness, metallic);
-#endif
-
   object_id = int(uint(resource_handle) & 0xFFFFu) + 1;
 }

source/blender/draw/engines/workbench/shaders/workbench_prepass_vert.glsl

@@ -1,44 +1,26 @@
 
+#pragma BLENDER_REQUIRE(common_view_clipping_lib.glsl)
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_shader_interface_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_material_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_image_lib.glsl)
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-in vec3 pos;
-in vec3 nor;
-in vec4 ac; /* active color */
-in vec2 au; /* active texture layer */
-#endif
-
 void main()
 {
   vec3 world_pos = point_object_to_world(pos);
   gl_Position = point_world_to_ndc(world_pos);
 
-#ifdef USE_WORLD_CLIP_PLANES
-  world_clip_planes_calc_clip_distance(world_pos);
-#endif
+  view_clipping_distances(world_pos);
 
   uv_interp = au;
 
   normal_interp = normalize(normal_object_to_view(nor));
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-#  ifdef OPAQUE_MATERIAL
-  float metallic, roughness;
-#  endif
-#endif
   workbench_material_data_get(resource_handle, color_interp, alpha_interp, roughness, metallic);
 
   if (materialIndex == 0) {
     color_interp = ac.rgb;
   }
 
-#ifdef OPAQUE_MATERIAL
-  packed_rough_metal = workbench_float_pair_encode(roughness, metallic);
-#endif
-
   object_id = int(uint(resource_handle) & 0xFFFFu) + 1;
 }

source/blender/draw/engines/workbench/shaders/workbench_shader_interface_lib.glsl

@@ -1,17 +0,0 @@
-
-#ifndef WORKBENCH_SHADER_SHARED_H
-IN_OUT ShaderStageInterface
-{
-  vec3 normal_interp;
-  vec3 color_interp;
-  float alpha_interp;
-  vec2 uv_interp;
-#  ifdef TRANSPARENT_MATERIAL
-  flat float roughness;
-  flat float metallic;
-#  else
-  flat float packed_rough_metal;
-#  endif
-  flat int object_id;
-};
-#endif

source/blender/draw/engines/workbench/shaders/workbench_shadow_caps_geom.glsl

@@ -2,38 +2,6 @@
 #  define USE_INVOC_EXT
 #endif
 
-#ifdef DOUBLE_MANIFOLD
-#  ifdef USE_INVOC_EXT
-#    define invoc_len 2
-#  else
-#    define vert_len 6
-#  endif
-#else
-#  ifdef USE_INVOC_EXT
-#    define invoc_len 2
-#  else
-#    define vert_len 6
-#  endif
-#endif
-
-#ifdef USE_INVOC_EXT
-layout(triangles, invocations = invoc_len) in;
-layout(triangle_strip, max_vertices = 3) out;
-#else
-layout(triangles) in;
-layout(triangle_strip, max_vertices = vert_len) out;
-#endif
-
-uniform vec3 lightDirection = vec3(0.57, 0.57, -0.57);
-
-in VertexData
-{
-  vec3 pos;           /* local position */
-  vec4 frontPosition; /* final ndc position */
-  vec4 backPosition;
-}
-vData[];
-
 vec4 get_pos(int v, bool backface)
 {
   return (backface) ? vData[v].backPosition : vData[v].frontPosition;
@@ -74,18 +42,19 @@ void main()
   /* In case of non manifold geom, we only increase/decrease
    * the stencil buffer by one but do every faces as they were facing the light. */
   bool invert = backface;
+  const bool is_manifold = false;
 #else
   const bool invert = false;
-  if (!backface) {
+  const bool is_manifold = true;
 #endif
+
+  if (!is_manifold || !backface) {
 #ifdef USE_INVOC_EXT
-  bool do_front = (gl_InvocationID & 1) == 0;
-  emit_cap(do_front, invert);
+    bool do_front = (gl_InvocationID & 1) == 0;
+    emit_cap(do_front, invert);
 #else
     emit_cap(true, invert);
     emit_cap(false, invert);
 #endif
-#ifndef DOUBLE_MANIFOLD
-}
-#endif
+  }
 }

source/blender/draw/engines/workbench/shaders/workbench_shadow_debug_frag.glsl

@@ -1,10 +1,4 @@
 
-out vec4 fragColor;
-
-layout(location = 0) out vec4 materialData;
-layout(location = 1) out vec4 normalData;
-layout(location = 2) out uint objectId;
-
 void main()
 {
   const float a = 0.25;

source/blender/draw/engines/workbench/shaders/workbench_shadow_geom.glsl

@@ -2,38 +2,6 @@
 #  define USE_INVOC_EXT
 #endif
 
-#ifdef DOUBLE_MANIFOLD
-#  ifdef USE_INVOC_EXT
-#    define invoc_len 2
-#  else
-#    define vert_len 8
-#  endif
-#else
-#  ifdef USE_INVOC_EXT
-#    define invoc_len 1
-#  else
-#    define vert_len 4
-#  endif
-#endif
-
-#ifdef USE_INVOC_EXT
-layout(lines_adjacency, invocations = invoc_len) in;
-layout(triangle_strip, max_vertices = 4) out;
-#else
-layout(lines_adjacency) in;
-layout(triangle_strip, max_vertices = vert_len) out;
-#endif
-
-uniform vec3 lightDirection = vec3(0.57, 0.57, -0.57);
-
-in VertexData
-{
-  vec3 pos;           /* local position */
-  vec4 frontPosition; /* final ndc position */
-  vec4 backPosition;
-}
-vData[];
-
 #define DEGENERATE_TRIS_WORKAROUND
 #define DEGENERATE_TRIS_AREA_THRESHOLD 4e-17
 

source/blender/draw/engines/workbench/shaders/workbench_shadow_vert.glsl

@@ -1,17 +1,5 @@
-#define INFINITE 1000.0
 
-uniform vec3 lightDirection = vec3(0.57, 0.57, -0.57);
-uniform float lightDistance = 1e4;
-
-in vec3 pos;
-
-out VertexData
-{
-  vec3 pos;           /* local position */
-  vec4 frontPosition; /* final ndc position */
-  vec4 backPosition;
-}
-vData;
+#pragma BLENDER_REQUIRE(common_view_lib.glsl)
 
 void main()
 {

source/blender/draw/engines/workbench/shaders/workbench_transparent_accum_frag.glsl

@@ -1,21 +1,10 @@
 
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_shader_interface_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_image_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_matcap_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_world_light_lib.glsl)
 
-#ifndef WORKBENCH_SHADER_SHARED_H
-/* Revealage is actually stored in transparentAccum alpha channel.
- * This is a workaround to older hardware not having separate blend equation per render target. */
-layout(location = 0) out vec4 transparentAccum;
-layout(location = 1) out vec4 revealageAccum;
-
-/* NOTE: Blending will be skipped on objectId because output is a non-normalized integer buffer. */
-layout(location = 2) out uint objectId;
-#endif
-
 /* Special function only to be used with calculate_transparent_weight(). */
 float linear_zdepth(float depth, vec4 viewvecs[2], mat4 proj_mat)
 {
@@ -69,7 +58,7 @@ void main()
 #endif
 
 #ifdef V3D_LIGHTING_MATCAP
-  vec3 shaded_color = get_matcap_lighting(color, N, I);
+  vec3 shaded_color = get_matcap_lighting(matcap_diffuse_tx, matcap_specular_tx, color, N, I);
 #endif
 
 #ifdef V3D_LIGHTING_STUDIO
@@ -80,7 +69,7 @@ void main()
   vec3 shaded_color = color;
 #endif
 
-  shaded_color *= get_shadow(N);
+  shaded_color *= get_shadow(N, forceShadowing);
 
   /* Listing 4 */
   float weight = calculate_transparent_weight() * alpha_interp;

source/blender/draw/engines/workbench/shaders/workbench_transparent_resolve_frag.glsl

@@ -1,11 +1,4 @@
 
-uniform sampler2D transparentAccum;
-uniform sampler2D transparentRevealage;
-
-in vec4 uvcoordsvar;
-
-out vec4 fragColor;
-
 /* Based on :
  * McGuire and Bavoil, Weighted Blended Order-Independent Transparency, Journal of
  * Computer Graphics Techniques (JCGT), vol. 2, no. 2, 122–141, 2013

source/blender/draw/engines/workbench/shaders/workbench_volume_frag.glsl

@@ -1,40 +1,8 @@
 
 #pragma BLENDER_REQUIRE(common_math_lib.glsl)
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(gpu_shader_common_obinfos_lib.glsl)
-#pragma BLENDER_REQUIRE(workbench_data_lib.glsl)
 #pragma BLENDER_REQUIRE(workbench_common_lib.glsl)
 
-uniform sampler2D depthBuffer;
-
-uniform sampler3D densityTexture;
-uniform sampler3D shadowTexture;
-uniform sampler3D flameTexture;
-uniform usampler3D flagTexture;
-uniform sampler1D flameColorTexture;
-uniform sampler1D transferTexture;
-uniform mat4 volumeObjectToTexture;
-
-uniform int samplesLen = 256;
-uniform float noiseOfs = 0.0;
-uniform float stepLength;   /* Step length in local space. */
-uniform float densityScale; /* Simple Opacity multiplicator. */
-uniform float gridScale;    /* Multiplicator for grid scaling. */
-uniform vec3 activeColor;
-
-uniform float slicePosition;
-uniform int sliceAxis; /* -1 is no slice, 0 is X, 1 is Y, 2 is Z. */
-
-uniform bool showPhi = false;
-uniform bool showFlags = false;
-uniform bool showPressure = false;
-
-#ifdef VOLUME_SLICE
-in vec3 localPos;
-#endif
-
-out vec4 fragColor;
-
 float phase_function_isotropic()
 {
   return 1.0 / (4.0 * M_PI);

source/blender/draw/engines/workbench/shaders/workbench_volume_vert.glsl

@@ -1,22 +1,8 @@
 
 #pragma BLENDER_REQUIRE(common_view_lib.glsl)
-#pragma BLENDER_REQUIRE(gpu_shader_common_obinfos_lib.glsl)
-
-uniform float slicePosition;
-uniform int sliceAxis; /* -1 is no slice, 0 is X, 1 is Y, 2 is Z. */
-
-uniform mat4 volumeTextureToObject;
-
-in vec3 pos;
 
 RESOURCE_ID_VARYING
 
-#ifdef VOLUME_SLICE
-in vec3 uvs;
-
-out vec3 localPos;
-#endif
-
 void main()
 {
 #ifdef VOLUME_SLICE

source/blender/draw/engines/workbench/shaders/workbench_world_light_lib.glsl

@@ -1,6 +1,4 @@
 
-#pragma BLENDER_REQUIRE(workbench_data_lib.glsl)
-
 /* [Drobot2014a] Low Level Optimizations for GCN */
 vec4 fast_rcp(vec4 v)
 {
@@ -120,12 +118,10 @@ vec3 get_world_lighting(vec3 base_color, float roughness, float metallic, vec3 N
   return diffuse_light + specular_light;
 }
 
-uniform bool forceShadowing = false;
-
-float get_shadow(vec3 N)
+float get_shadow(vec3 N, bool force_shadowing)
 {
   float light_factor = -dot(N, world_data.shadow_direction_vs.xyz);
   float shadow_mix = smoothstep(world_data.shadow_shift, world_data.shadow_focus, light_factor);
-  shadow_mix *= forceShadowing ? 0.0 : world_data.shadow_mul;
+  shadow_mix *= force_shadowing ? 0.0 : world_data.shadow_mul;
   return shadow_mix + world_data.shadow_add;
 }

source/blender/draw/engines/workbench/workbench_effect_cavity.c

@@ -165,7 +165,7 @@ void workbench_cavity_cache_init(WORKBENCH_Data *data)
     grp = DRW_shgroup_create(sh, psl->cavity_ps);
     DRW_shgroup_uniform_texture(grp, "normalBuffer", wpd->normal_buffer_tx);
     DRW_shgroup_uniform_block(grp, "samples_block", wpd->vldata->cavity_sample_ubo);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
 
     if (SSAO_ENABLED(wpd)) {
       DRW_shgroup_uniform_texture(grp, "depthBuffer", dtxl->depth);

source/blender/draw/engines/workbench/workbench_effect_dof.c

@@ -328,7 +328,7 @@ void workbench_dof_cache_init(WORKBENCH_Data *vedata)
 
     float offset = wpd->taa_sample / (float)max_ii(1, wpd->taa_sample_len);
     DRWShadingGroup *grp = DRW_shgroup_create(blur1_sh, psl->dof_blur1_ps);
-    DRW_shgroup_uniform_block(grp, "dofSamplesBlock", wpd->vldata->dof_sample_ubo);
+    DRW_shgroup_uniform_block(grp, "samples", wpd->vldata->dof_sample_ubo);
     DRW_shgroup_uniform_texture(grp, "noiseTex", wpd->vldata->cavity_jitter_tx);
     DRW_shgroup_uniform_texture(grp, "inputCocTex", txl->coc_halfres_tx);
     DRW_shgroup_uniform_texture(grp, "halfResColorTex", txl->dof_source_tx);

source/blender/draw/engines/workbench/workbench_effect_outline.c

@@ -46,7 +46,7 @@ void workbench_outline_cache_init(WORKBENCH_Data *data)
     grp = DRW_shgroup_create(sh, psl->outline_ps);
     DRW_shgroup_uniform_texture(grp, "objectIdBuffer", wpd->object_id_tx);
     DRW_shgroup_uniform_texture(grp, "depthBuffer", dtxl->depth);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
     DRW_shgroup_call_procedural_triangles(grp, NULL, 1);
   }
   else {

source/blender/draw/engines/workbench/workbench_engine.c

@@ -51,8 +51,6 @@ void workbench_engine_init(void *ved)
   WORKBENCH_StorageList *stl = vedata->stl;
   WORKBENCH_TextureList *txl = vedata->txl;
 
-  workbench_shader_library_ensure();
-
   workbench_private_data_alloc(stl);
   WORKBENCH_PrivateData *wpd = stl->wpd;
   workbench_private_data_init(wpd);

source/blender/draw/engines/workbench/workbench_materials.c

@@ -210,7 +210,7 @@ DRWShadingGroup *workbench_material_setup_ex(WORKBENCH_PrivateData *wpd,
 
       DRWShadingGroup *grp = prepass->common_shgrp;
       *grp_mat = grp = DRW_shgroup_create_sub(grp);
-      DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+      DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
       DRW_shgroup_uniform_int_copy(grp, "materialIndex", mat_id);
       return grp;
     }
@@ -234,7 +234,7 @@ DRWShadingGroup *workbench_material_setup_ex(WORKBENCH_PrivateData *wpd,
       DRWShadingGroup **grp = &wpd->prepass[transp][infront][datatype].common_shgrp;
       if (resource_changed) {
         *grp = DRW_shgroup_create_sub(*grp);
-        DRW_shgroup_uniform_block(*grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(*grp, "materials_data", wpd->material_ubo_curr);
       }
       if (r_transp && transp) {
         *r_transp = true;
@@ -293,5 +293,6 @@ DRWShadingGroup *workbench_image_setup_ex(WORKBENCH_PrivateData *wpd,
     DRW_shgroup_uniform_texture_ex(grp, "imageTexture", tex, sampler);
   }
   DRW_shgroup_uniform_bool_copy(grp, "imagePremult", (ima && ima->alpha_mode == IMA_ALPHA_PREMUL));
+  DRW_shgroup_uniform_float_copy(grp, "imageTransparencyCutoff", 0.1f);
   return grp;
 }

source/blender/draw/engines/workbench/workbench_opaque.c

@@ -88,26 +88,26 @@ void workbench_opaque_cache_init(WORKBENCH_Data *vedata)
         sh = workbench_shader_opaque_get(wpd, data);
 
         wpd->prepass[opaque][infront][data].common_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", -1);
         DRW_shgroup_uniform_bool_copy(grp, "useMatcap", use_matcap);
 
         wpd->prepass[opaque][infront][data].vcol_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", 0); /* Default material. (uses vcol) */
         DRW_shgroup_uniform_bool_copy(grp, "useMatcap", use_matcap);
 
         sh = workbench_shader_opaque_image_get(wpd, data, false);
 
         wpd->prepass[opaque][infront][data].image_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", 0); /* Default material. */
         DRW_shgroup_uniform_bool_copy(grp, "useMatcap", use_matcap);
 
         sh = workbench_shader_opaque_image_get(wpd, data, true);
 
         wpd->prepass[opaque][infront][data].image_tiled_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", 0); /* Default material. */
         DRW_shgroup_uniform_bool_copy(grp, "useMatcap", use_matcap);
       }
@@ -121,7 +121,7 @@ void workbench_opaque_cache_init(WORKBENCH_Data *vedata)
     sh = workbench_shader_composite_get(wpd);
 
     grp = DRW_shgroup_create(sh, psl->composite_ps);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
     DRW_shgroup_uniform_texture(grp, "materialBuffer", wpd->material_buffer_tx);
     DRW_shgroup_uniform_texture(grp, "normalBuffer", wpd->normal_buffer_tx);
     DRW_shgroup_uniform_bool_copy(grp, "forceShadowing", false);
@@ -135,8 +135,8 @@ void workbench_opaque_cache_init(WORKBENCH_Data *vedata)
       struct GPUTexture *spec_tx = wpd->studio_light->matcap_specular.gputexture;
       const bool use_spec = workbench_is_specular_highlight_enabled(wpd);
       spec_tx = (use_spec && spec_tx) ? spec_tx : diff_tx;
-      DRW_shgroup_uniform_texture(grp, "matcapDiffuseImage", diff_tx);
-      DRW_shgroup_uniform_texture(grp, "matcapSpecularImage", spec_tx);
+      DRW_shgroup_uniform_texture(grp, "matcap_diffuse_tx", diff_tx);
+      DRW_shgroup_uniform_texture(grp, "matcap_specular_tx", spec_tx);
     }
     DRW_shgroup_call_procedural_triangles(grp, NULL, 1);
 

source/blender/draw/engines/workbench/workbench_private.h

@@ -459,7 +459,6 @@ void workbench_shader_depth_of_field_get(GPUShader **prepare_sh,
                                          GPUShader **blur2_sh,
                                          GPUShader **resolve_sh);
 
-void workbench_shader_library_ensure(void);
 void workbench_shader_free(void);
 
 /* workbench_effect_antialiasing.c */

source/blender/draw/engines/workbench/workbench_shader.c

@@ -1,562 +0,0 @@
-/*
- * 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.
- *
- * Copyright 2020, Blender Foundation.
- */
-
-/** \file
- * \ingroup draw_engine
- */
-
-#include "DRW_render.h"
-
-#include "BLI_dynstr.h"
-#include "BLI_string_utils.h"
-
-#include "workbench_engine.h"
-#include "workbench_private.h"
-
-extern char datatoc_common_math_lib_glsl[];
-extern char datatoc_common_math_geom_lib_glsl[];
-extern char datatoc_common_hair_lib_glsl[];
-extern char datatoc_common_pointcloud_lib_glsl[];
-extern char datatoc_common_view_lib_glsl[];
-extern char datatoc_common_smaa_lib_glsl[];
-
-extern char datatoc_workbench_prepass_vert_glsl[];
-extern char datatoc_workbench_prepass_hair_vert_glsl[];
-extern char datatoc_workbench_prepass_pointcloud_vert_glsl[];
-extern char datatoc_workbench_prepass_frag_glsl[];
-
-extern char datatoc_workbench_effect_cavity_frag_glsl[];
-extern char datatoc_workbench_effect_outline_frag_glsl[];
-extern char datatoc_workbench_effect_dof_frag_glsl[];
-extern char datatoc_workbench_effect_taa_frag_glsl[];
-extern char datatoc_workbench_effect_smaa_frag_glsl[];
-extern char datatoc_workbench_effect_smaa_vert_glsl[];
-
-extern char datatoc_workbench_composite_frag_glsl[];
-
-extern char datatoc_workbench_transparent_accum_frag_glsl[];
-extern char datatoc_workbench_transparent_resolve_frag_glsl[];
-
-extern char datatoc_workbench_merge_infront_frag_glsl[];
-
-extern char datatoc_workbench_shadow_vert_glsl[];
-extern char datatoc_workbench_shadow_geom_glsl[];
-extern char datatoc_workbench_shadow_caps_geom_glsl[];
-extern char datatoc_workbench_shadow_debug_frag_glsl[];
-
-extern char datatoc_workbench_volume_vert_glsl[];
-extern char datatoc_workbench_volume_frag_glsl[];
-
-extern char datatoc_workbench_cavity_lib_glsl[];
-extern char datatoc_workbench_common_lib_glsl[];
-extern char datatoc_workbench_curvature_lib_glsl[];
-extern char datatoc_workbench_data_lib_glsl[];
-extern char datatoc_workbench_image_lib_glsl[];
-extern char datatoc_workbench_matcap_lib_glsl[];
-extern char datatoc_workbench_material_lib_glsl[];
-extern char datatoc_workbench_shader_interface_lib_glsl[];
-extern char datatoc_workbench_world_light_lib_glsl[];
-
-extern char datatoc_gpu_shader_depth_only_frag_glsl[];
-extern char datatoc_gpu_shader_common_obinfos_lib_glsl[];
-
-/* Maximum number of variations. */
-#define MAX_LIGHTING 3
-#define MAX_COLOR 3
-
-enum {
-  VOLUME_SH_SLICE = 0,
-  VOLUME_SH_COBA,
-  VOLUME_SH_CUBIC,
-};
-
-#define VOLUME_SH_MAX (1 << (VOLUME_SH_CUBIC + 1))
-
-static struct {
-  struct GPUShader *opaque_prepass_sh_cache[GPU_SHADER_CFG_LEN][WORKBENCH_DATATYPE_MAX][MAX_COLOR];
-  struct GPUShader *transp_prepass_sh_cache[GPU_SHADER_CFG_LEN][WORKBENCH_DATATYPE_MAX]
-                                           [MAX_LIGHTING][MAX_COLOR];
-
-  struct GPUShader *opaque_composite_sh[MAX_LIGHTING];
-  struct GPUShader *oit_resolve_sh;
-  struct GPUShader *outline_sh;
-  struct GPUShader *merge_infront_sh;
-
-  struct GPUShader *shadow_depth_pass_sh[2];
-  struct GPUShader *shadow_depth_fail_sh[2][2];
-
-  struct GPUShader *cavity_sh[2][2];
-
-  struct GPUShader *dof_prepare_sh;
-  struct GPUShader *dof_downsample_sh;
-  struct GPUShader *dof_blur1_sh;
-  struct GPUShader *dof_blur2_sh;
-  struct GPUShader *dof_resolve_sh;
-
-  struct GPUShader *aa_accum_sh;
-  struct GPUShader *smaa_sh[3];
-
-  struct GPUShader *volume_sh[2][2][3][2];
-
-  struct DRWShaderLibrary *lib;
-} e_data = {{{{NULL}}}};
-
-void workbench_shader_library_ensure(void)
-{
-  if (e_data.lib == NULL) {
-    e_data.lib = DRW_shader_library_create();
-    /* NOTE: These need to be ordered by dependencies. */
-    DRW_SHADER_LIB_ADD(e_data.lib, common_math_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, common_math_geom_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, common_hair_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, common_view_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, common_pointcloud_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, gpu_shader_common_obinfos_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_shader_interface_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_common_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_image_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_material_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_data_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_matcap_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_cavity_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_curvature_lib);
-    DRW_SHADER_LIB_ADD(e_data.lib, workbench_world_light_lib);
-  }
-}
-
-static char *workbench_build_defines(
-    WORKBENCH_PrivateData *wpd, bool textured, bool tiled, bool cavity, bool curvature)
-{
-  char *str = NULL;
-
-  DynStr *ds = BLI_dynstr_new();
-
-  if (wpd && wpd->shading.light == V3D_LIGHTING_STUDIO) {
-    BLI_dynstr_append(ds, "#define V3D_LIGHTING_STUDIO\n");
-  }
-  else if (wpd && wpd->shading.light == V3D_LIGHTING_MATCAP) {
-    BLI_dynstr_append(ds, "#define V3D_LIGHTING_MATCAP\n");
-  }
-  else {
-    BLI_dynstr_append(ds, "#define V3D_LIGHTING_FLAT\n");
-  }
-
-  if (NORMAL_ENCODING_ENABLED()) {
-    BLI_dynstr_append(ds, "#define WORKBENCH_ENCODE_NORMALS\n");
-  }
-
-  if (textured) {
-    BLI_dynstr_append(ds, "#define V3D_SHADING_TEXTURE_COLOR\n");
-  }
-  if (tiled) {
-    BLI_dynstr_append(ds, "#define TEXTURE_IMAGE_ARRAY\n");
-  }
-  if (cavity) {
-    BLI_dynstr_append(ds, "#define USE_CAVITY\n");
-  }
-  if (curvature) {
-    BLI_dynstr_append(ds, "#define USE_CURVATURE\n");
-  }
-
-  str = BLI_dynstr_get_cstring(ds);
-  BLI_dynstr_free(ds);
-  return str;
-}
-
-static int workbench_color_index(WORKBENCH_PrivateData *UNUSED(wpd), bool textured, bool tiled)
-{
-  BLI_assert(2 < MAX_COLOR);
-  return (textured) ? (tiled ? 2 : 1) : 0;
-}
-
-static GPUShader *workbench_shader_get_ex(WORKBENCH_PrivateData *wpd,
-                                          bool transp,
-                                          eWORKBENCH_DataType datatype,
-                                          bool textured,
-                                          bool tiled)
-{
-  int color = workbench_color_index(wpd, textured, tiled);
-  int light = wpd->shading.light;
-  BLI_assert(light < MAX_LIGHTING);
-  struct GPUShader **shader =
-      (transp) ? &e_data.transp_prepass_sh_cache[wpd->sh_cfg][datatype][light][color] :
-                 &e_data.opaque_prepass_sh_cache[wpd->sh_cfg][datatype][color];
-
-  if (*shader == NULL) {
-    char *defines = workbench_build_defines(wpd, textured, tiled, false, false);
-
-    char *frag_file = transp ? datatoc_workbench_transparent_accum_frag_glsl :
-                               datatoc_workbench_prepass_frag_glsl;
-    char *frag_src = DRW_shader_library_create_shader_string(e_data.lib, frag_file);
-
-    char *vert_file = (datatype == WORKBENCH_DATATYPE_HAIR) ?
-                          datatoc_workbench_prepass_hair_vert_glsl :
-                          ((datatype == WORKBENCH_DATATYPE_POINTCLOUD) ?
-                               datatoc_workbench_prepass_pointcloud_vert_glsl :
-                               datatoc_workbench_prepass_vert_glsl);
-    char *vert_src = DRW_shader_library_create_shader_string(e_data.lib, vert_file);
-
-    const GPUShaderConfigData *sh_cfg_data = &GPU_shader_cfg_data[wpd->sh_cfg];
-
-    *shader = GPU_shader_create_from_arrays({
-        .vert = (const char *[]){sh_cfg_data->lib, vert_src, NULL},
-        .frag = (const char *[]){frag_src, NULL},
-        .defs = (const char *[]){sh_cfg_data->def,
-                                 defines,
-                                 transp ? "#define TRANSPARENT_MATERIAL\n" :
-                                          "#define OPAQUE_MATERIAL\n",
-                                 (datatype == WORKBENCH_DATATYPE_POINTCLOUD) ?
-                                     "#define UNIFORM_RESOURCE_ID\n"
-                                     "#define INSTANCED_ATTR\n" :
-                                     NULL,
-                                 NULL},
-    });
-
-    MEM_freeN(defines);
-    MEM_freeN(frag_src);
-    MEM_freeN(vert_src);
-  }
-  return *shader;
-}
-
-GPUShader *workbench_shader_opaque_get(WORKBENCH_PrivateData *wpd, eWORKBENCH_DataType datatype)
-{
-  return workbench_shader_get_ex(wpd, false, datatype, false, false);
-}
-
-GPUShader *workbench_shader_opaque_image_get(WORKBENCH_PrivateData *wpd,
-                                             eWORKBENCH_DataType datatype,
-                                             bool tiled)
-{
-  return workbench_shader_get_ex(wpd, false, datatype, true, tiled);
-}
-
-GPUShader *workbench_shader_transparent_get(WORKBENCH_PrivateData *wpd,
-                                            eWORKBENCH_DataType datatype)
-{
-  return workbench_shader_get_ex(wpd, true, datatype, false, false);
-}
-
-GPUShader *workbench_shader_transparent_image_get(WORKBENCH_PrivateData *wpd,
-                                                  eWORKBENCH_DataType datatype,
-                                                  bool tiled)
-{
-  return workbench_shader_get_ex(wpd, true, datatype, true, tiled);
-}
-
-GPUShader *workbench_shader_composite_get(WORKBENCH_PrivateData *wpd)
-{
-  int light = wpd->shading.light;
-  struct GPUShader **shader = &e_data.opaque_composite_sh[light];
-  BLI_assert(light < MAX_LIGHTING);
-
-  if (*shader == NULL) {
-    char *defines = workbench_build_defines(wpd, false, false, false, false);
-    char *frag = DRW_shader_library_create_shader_string(e_data.lib,
-                                                         datatoc_workbench_composite_frag_glsl);
-
-    *shader = DRW_shader_create_fullscreen(frag, defines);
-
-    MEM_freeN(defines);
-    MEM_freeN(frag);
-  }
-  return *shader;
-}
-
-GPUShader *workbench_shader_merge_infront_get(WORKBENCH_PrivateData *UNUSED(wpd))
-{
-  if (e_data.merge_infront_sh == NULL) {
-    char *frag = DRW_shader_library_create_shader_string(
-        e_data.lib, datatoc_workbench_merge_infront_frag_glsl);
-
-    e_data.merge_infront_sh = DRW_shader_create_fullscreen(frag, NULL);
-
-    MEM_freeN(frag);
-  }
-  return e_data.merge_infront_sh;
-}
-
-GPUShader *workbench_shader_transparent_resolve_get(WORKBENCH_PrivateData *wpd)
-{
-  if (e_data.oit_resolve_sh == NULL) {
-    char *defines = workbench_build_defines(wpd, false, false, false, false);
-
-    e_data.oit_resolve_sh = DRW_shader_create_fullscreen(
-        datatoc_workbench_transparent_resolve_frag_glsl, defines);
-
-    MEM_freeN(defines);
-  }
-  return e_data.oit_resolve_sh;
-}
-
-static GPUShader *workbench_shader_shadow_pass_get_ex(bool depth_pass, bool manifold, bool cap)
-{
-  struct GPUShader **shader = (depth_pass) ? &e_data.shadow_depth_pass_sh[manifold] :
-                                             &e_data.shadow_depth_fail_sh[manifold][cap];
-
-  if (*shader == NULL) {
-#if DEBUG_SHADOW_VOLUME
-    const char *shadow_frag = datatoc_workbench_shadow_debug_frag_glsl;
-#else
-    const char *shadow_frag = datatoc_gpu_shader_depth_only_frag_glsl;
-#endif
-
-    *shader = GPU_shader_create_from_arrays({
-        .vert = (const char *[]){datatoc_common_view_lib_glsl,
-                                 datatoc_workbench_shadow_vert_glsl,
-                                 NULL},
-        .geom = (const char *[]){(cap) ? datatoc_workbench_shadow_caps_geom_glsl :
-                                         datatoc_workbench_shadow_geom_glsl,
-                                 NULL},
-        .frag = (const char *[]){shadow_frag, NULL},
-        .defs = (const char *[]){(depth_pass) ? "#define SHADOW_PASS\n" : "#define SHADOW_FAIL\n",
-                                 (manifold) ? "" : "#define DOUBLE_MANIFOLD\n",
-                                 NULL},
-    });
-  }
-  return *shader;
-}
-
-GPUShader *workbench_shader_shadow_pass_get(bool manifold)
-{
-  return workbench_shader_shadow_pass_get_ex(true, manifold, false);
-}
-
-GPUShader *workbench_shader_shadow_fail_get(bool manifold, bool cap)
-{
-  return workbench_shader_shadow_pass_get_ex(false, manifold, cap);
-}
-
-GPUShader *workbench_shader_cavity_get(bool cavity, bool curvature)
-{
-  BLI_assert(cavity || curvature);
-  struct GPUShader **shader = &e_data.cavity_sh[cavity][curvature];
-
-  if (*shader == NULL) {
-    char *defines = workbench_build_defines(NULL, false, false, cavity, curvature);
-    char *frag = DRW_shader_library_create_shader_string(
-        e_data.lib, datatoc_workbench_effect_cavity_frag_glsl);
-
-    *shader = DRW_shader_create_fullscreen(frag, defines);
-
-    MEM_freeN(defines);
-    MEM_freeN(frag);
-  }
-  return *shader;
-}
-
-GPUShader *workbench_shader_outline_get(void)
-{
-  if (e_data.outline_sh == NULL) {
-    char *frag = DRW_shader_library_create_shader_string(
-        e_data.lib, datatoc_workbench_effect_outline_frag_glsl);
-
-    e_data.outline_sh = DRW_shader_create_fullscreen(frag, NULL);
-
-    MEM_freeN(frag);
-  }
-  return e_data.outline_sh;
-}
-
-void workbench_shader_depth_of_field_get(GPUShader **prepare_sh,
-                                         GPUShader **downsample_sh,
-                                         GPUShader **blur1_sh,
-                                         GPUShader **blur2_sh,
-                                         GPUShader **resolve_sh)
-{
-  if (e_data.dof_prepare_sh == NULL) {
-    e_data.dof_prepare_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define PREPARE\n");
-    e_data.dof_downsample_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define DOWNSAMPLE\n");
-#if 0 /* TODO(fclem): finish COC min_max optimization */
-    e_data.dof_flatten_v_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define FLATTEN_VERTICAL\n");
-    e_data.dof_flatten_h_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define FLATTEN_HORIZONTAL\n");
-    e_data.dof_dilate_v_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define DILATE_VERTICAL\n");
-    e_data.dof_dilate_h_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define DILATE_HORIZONTAL\n");
-#endif
-    e_data.dof_blur1_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define BLUR1\n");
-    e_data.dof_blur2_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define BLUR2\n");
-    e_data.dof_resolve_sh = DRW_shader_create_fullscreen_with_shaderlib(
-        datatoc_workbench_effect_dof_frag_glsl, e_data.lib, "#define RESOLVE\n");
-  }
-
-  *prepare_sh = e_data.dof_prepare_sh;
-  *downsample_sh = e_data.dof_downsample_sh;
-  *blur1_sh = e_data.dof_blur1_sh;
-  *blur2_sh = e_data.dof_blur2_sh;
-  *resolve_sh = e_data.dof_resolve_sh;
-}
-
-GPUShader *workbench_shader_antialiasing_accumulation_get(void)
-{
-  if (e_data.aa_accum_sh == NULL) {
-    char *frag = DRW_shader_library_create_shader_string(e_data.lib,
-                                                         datatoc_workbench_effect_taa_frag_glsl);
-
-    e_data.aa_accum_sh = DRW_shader_create_fullscreen(frag, NULL);
-
-    MEM_freeN(frag);
-  }
-  return e_data.aa_accum_sh;
-}
-
-GPUShader *workbench_shader_antialiasing_get(int stage)
-{
-  BLI_assert(stage < 3);
-  if (!e_data.smaa_sh[stage]) {
-    char stage_define[32];
-    BLI_snprintf(stage_define, sizeof(stage_define), "#define SMAA_STAGE %d\n", stage);
-
-    e_data.smaa_sh[stage] = GPU_shader_create_from_arrays({
-        .vert =
-            (const char *[]){
-                "#define SMAA_INCLUDE_VS 1\n",
-                "#define SMAA_INCLUDE_PS 0\n",
-                "uniform vec4 viewportMetrics;\n",
-                datatoc_common_smaa_lib_glsl,
-                datatoc_workbench_effect_smaa_vert_glsl,
-                NULL,
-            },
-        .frag =
-            (const char *[]){
-                "#define SMAA_INCLUDE_VS 0\n",
-                "#define SMAA_INCLUDE_PS 1\n",
-                "uniform vec4 viewportMetrics;\n",
-                datatoc_common_smaa_lib_glsl,
-                datatoc_workbench_effect_smaa_frag_glsl,
-                NULL,
-            },
-        .defs =
-            (const char *[]){
-                "#define SMAA_GLSL_3\n",
-                "#define SMAA_RT_METRICS viewportMetrics\n",
-                "#define SMAA_PRESET_HIGH\n",
-                "#define SMAA_LUMA_WEIGHT float4(1.0, 1.0, 1.0, 1.0)\n",
-                "#define SMAA_NO_DISCARD\n",
-                stage_define,
-                NULL,
-            },
-    });
-  }
-  return e_data.smaa_sh[stage];
-}
-
-GPUShader *workbench_shader_volume_get(bool slice,
-                                       bool coba,
-                                       eWORKBENCH_VolumeInterpType interp_type,
-                                       bool smoke)
-{
-  GPUShader **shader = &e_data.volume_sh[slice][coba][interp_type][smoke];
-
-  if (*shader == NULL) {
-    DynStr *ds = BLI_dynstr_new();
-
-    if (slice) {
-      BLI_dynstr_append(ds, "#define VOLUME_SLICE\n");
-    }
-    if (coba) {
-      BLI_dynstr_append(ds, "#define USE_COBA\n");
-    }
-    switch (interp_type) {
-      case WORKBENCH_VOLUME_INTERP_LINEAR:
-        BLI_dynstr_append(ds, "#define USE_TRILINEAR\n");
-        break;
-      case WORKBENCH_VOLUME_INTERP_CUBIC:
-        BLI_dynstr_append(ds, "#define USE_TRICUBIC\n");
-        break;
-      case WORKBENCH_VOLUME_INTERP_CLOSEST:
-        BLI_dynstr_append(ds, "#define USE_CLOSEST\n");
-        break;
-    }
-    if (smoke) {
-      BLI_dynstr_append(ds, "#define VOLUME_SMOKE\n");
-    }
-
-    char *defines = BLI_dynstr_get_cstring(ds);
-    BLI_dynstr_free(ds);
-
-    char *vert = DRW_shader_library_create_shader_string(e_data.lib,
-                                                         datatoc_workbench_volume_vert_glsl);
-    char *frag = DRW_shader_library_create_shader_string(e_data.lib,
-                                                         datatoc_workbench_volume_frag_glsl);
-
-    *shader = DRW_shader_create(vert, NULL, frag, defines);
-
-    MEM_freeN(vert);
-    MEM_freeN(frag);
-    MEM_freeN(defines);
-  }
-  return *shader;
-}
-
-void workbench_shader_free(void)
-{
-  for (int j = 0; j < sizeof(e_data.opaque_prepass_sh_cache) / sizeof(void *); j++) {
-    struct GPUShader **sh_array = &e_data.opaque_prepass_sh_cache[0][0][0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < sizeof(e_data.transp_prepass_sh_cache) / sizeof(void *); j++) {
-    struct GPUShader **sh_array = &e_data.transp_prepass_sh_cache[0][0][0][0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < ARRAY_SIZE(e_data.opaque_composite_sh); j++) {
-    struct GPUShader **sh_array = &e_data.opaque_composite_sh[0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < ARRAY_SIZE(e_data.shadow_depth_pass_sh); j++) {
-    struct GPUShader **sh_array = &e_data.shadow_depth_pass_sh[0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < sizeof(e_data.shadow_depth_fail_sh) / sizeof(void *); j++) {
-    struct GPUShader **sh_array = &e_data.shadow_depth_fail_sh[0][0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < sizeof(e_data.cavity_sh) / sizeof(void *); j++) {
-    struct GPUShader **sh_array = &e_data.cavity_sh[0][0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < ARRAY_SIZE(e_data.smaa_sh); j++) {
-    struct GPUShader **sh_array = &e_data.smaa_sh[0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-  for (int j = 0; j < sizeof(e_data.volume_sh) / sizeof(void *); j++) {
-    struct GPUShader **sh_array = &e_data.volume_sh[0][0][0][0];
-    DRW_SHADER_FREE_SAFE(sh_array[j]);
-  }
-
-  DRW_SHADER_FREE_SAFE(e_data.oit_resolve_sh);
-  DRW_SHADER_FREE_SAFE(e_data.outline_sh);
-  DRW_SHADER_FREE_SAFE(e_data.merge_infront_sh);
-
-  DRW_SHADER_FREE_SAFE(e_data.dof_prepare_sh);
-  DRW_SHADER_FREE_SAFE(e_data.dof_downsample_sh);
-  DRW_SHADER_FREE_SAFE(e_data.dof_blur1_sh);
-  DRW_SHADER_FREE_SAFE(e_data.dof_blur2_sh);
-  DRW_SHADER_FREE_SAFE(e_data.dof_resolve_sh);
-
-  DRW_SHADER_FREE_SAFE(e_data.aa_accum_sh);
-
-  DRW_SHADER_LIB_FREE_SAFE(e_data.lib);
-}

source/blender/draw/engines/workbench/workbench_shader.cc

@@ -0,0 +1,398 @@
+/*
+ * 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.
+ *
+ * Copyright 2020, Blender Foundation.
+ */
+
+/** \file
+ * \ingroup draw_engine
+ */
+
+#include "DRW_render.h"
+
+#include <string>
+
+#include "workbench_engine.h"
+#include "workbench_private.h"
+
+/* Maximum number of variations. */
+#define MAX_LIGHTING 3
+
+enum eWORKBENCH_TextureType {
+  TEXTURE_SH_NONE = 0,
+  TEXTURE_SH_SINGLE,
+  TEXTURE_SH_TILED,
+  TEXTURE_SH_MAX,
+};
+
+static struct {
+  struct GPUShader
+      *opaque_prepass_sh_cache[GPU_SHADER_CFG_LEN][WORKBENCH_DATATYPE_MAX][TEXTURE_SH_MAX];
+  struct GPUShader *transp_prepass_sh_cache[GPU_SHADER_CFG_LEN][WORKBENCH_DATATYPE_MAX]
+                                           [MAX_LIGHTING][TEXTURE_SH_MAX];
+
+  struct GPUShader *opaque_composite_sh[MAX_LIGHTING];
+  struct GPUShader *oit_resolve_sh;
+  struct GPUShader *outline_sh;
+  struct GPUShader *merge_infront_sh;
+
+  struct GPUShader *shadow_depth_pass_sh[2];
+  struct GPUShader *shadow_depth_fail_sh[2][2];
+
+  struct GPUShader *cavity_sh[2][2];
+
+  struct GPUShader *dof_prepare_sh;
+  struct GPUShader *dof_downsample_sh;
+  struct GPUShader *dof_blur1_sh;
+  struct GPUShader *dof_blur2_sh;
+  struct GPUShader *dof_resolve_sh;
+
+  struct GPUShader *aa_accum_sh;
+  struct GPUShader *smaa_sh[3];
+
+  struct GPUShader *volume_sh[2][2][3][2];
+
+} e_data = {{{{NULL}}}};
+
+/* -------------------------------------------------------------------- */
+/** \name Conversions
+ * \{ */
+
+static const char *workbench_lighting_mode_to_str(int light)
+{
+  switch (light) {
+    default:
+      BLI_assert_msg(0, "Error: Unknown lighting mode.");
+      ATTR_FALLTHROUGH;
+    case V3D_LIGHTING_STUDIO:
+      return "_studio";
+    case V3D_LIGHTING_MATCAP:
+      return "_matcap";
+    case V3D_LIGHTING_FLAT:
+      return "_flat";
+      return "";
+  }
+}
+
+static const char *workbench_datatype_mode_to_str(eWORKBENCH_DataType datatype)
+{
+  switch (datatype) {
+    default:
+      BLI_assert_msg(0, "Error: Unknown data mode.");
+      ATTR_FALLTHROUGH;
+    case WORKBENCH_DATATYPE_MESH:
+      return "_mesh";
+    case WORKBENCH_DATATYPE_HAIR:
+      return "_hair";
+    case WORKBENCH_DATATYPE_POINTCLOUD:
+      return "_ptcloud";
+  }
+}
+
+static const char *workbench_volume_interp_to_str(eWORKBENCH_VolumeInterpType interp_type)
+{
+  switch (interp_type) {
+    default:
+      BLI_assert_msg(0, "Error: Unknown lighting mode.");
+      ATTR_FALLTHROUGH;
+    case WORKBENCH_VOLUME_INTERP_LINEAR:
+      return "_linear";
+    case WORKBENCH_VOLUME_INTERP_CUBIC:
+      return "_cubic";
+    case WORKBENCH_VOLUME_INTERP_CLOSEST:
+      return "_closest";
+  }
+}
+
+static const char *workbench_texture_type_to_str(eWORKBENCH_TextureType tex_type)
+{
+  switch (tex_type) {
+    default:
+      BLI_assert_msg(0, "Error: Unknown texture mode.");
+      ATTR_FALLTHROUGH;
+    case TEXTURE_SH_NONE:
+      return "_tex_none";
+    case TEXTURE_SH_TILED:
+      return "_tex_tile";
+    case TEXTURE_SH_SINGLE:
+      return "_tex_single";
+  }
+}
+
+static eWORKBENCH_TextureType workbench_texture_type_get(bool textured, bool tiled)
+{
+  return textured ? (tiled ? TEXTURE_SH_TILED : TEXTURE_SH_SINGLE) : TEXTURE_SH_NONE;
+}
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Shader request
+ * \{ */
+
+static GPUShader *workbench_shader_get_ex(WORKBENCH_PrivateData *wpd,
+                                          bool transp,
+                                          eWORKBENCH_DataType datatype,
+                                          bool textured,
+                                          bool tiled)
+{
+  eWORKBENCH_TextureType tex_type = workbench_texture_type_get(textured, tiled);
+  int light = wpd->shading.light;
+  BLI_assert(light < MAX_LIGHTING);
+  struct GPUShader **shader =
+      (transp) ? &e_data.transp_prepass_sh_cache[wpd->sh_cfg][datatype][light][tex_type] :
+                 &e_data.opaque_prepass_sh_cache[wpd->sh_cfg][datatype][tex_type];
+
+  if (*shader == nullptr) {
+    std::string create_info_name = "workbench";
+    create_info_name += (transp) ? "_transp" : "_opaque";
+    if (transp) {
+      create_info_name += workbench_lighting_mode_to_str(light);
+    }
+    create_info_name += workbench_datatype_mode_to_str(datatype);
+    create_info_name += workbench_texture_type_to_str(tex_type);
+    create_info_name += (wpd->sh_cfg == GPU_SHADER_CFG_CLIPPED) ? "_clip" : "_no_clip";
+
+    *shader = GPU_shader_create_from_info_name(create_info_name.c_str());
+  }
+  return *shader;
+}
+
+GPUShader *workbench_shader_opaque_get(WORKBENCH_PrivateData *wpd, eWORKBENCH_DataType datatype)
+{
+  return workbench_shader_get_ex(wpd, false, datatype, false, false);
+}
+
+GPUShader *workbench_shader_opaque_image_get(WORKBENCH_PrivateData *wpd,
+                                             eWORKBENCH_DataType datatype,
+                                             bool tiled)
+{
+  return workbench_shader_get_ex(wpd, false, datatype, true, tiled);
+}
+
+GPUShader *workbench_shader_transparent_get(WORKBENCH_PrivateData *wpd,
+                                            eWORKBENCH_DataType datatype)
+{
+  return workbench_shader_get_ex(wpd, true, datatype, false, false);
+}
+
+GPUShader *workbench_shader_transparent_image_get(WORKBENCH_PrivateData *wpd,
+                                                  eWORKBENCH_DataType datatype,
+                                                  bool tiled)
+{
+  return workbench_shader_get_ex(wpd, true, datatype, true, tiled);
+}
+
+GPUShader *workbench_shader_composite_get(WORKBENCH_PrivateData *wpd)
+{
+  int light = wpd->shading.light;
+  struct GPUShader **shader = &e_data.opaque_composite_sh[light];
+  BLI_assert(light < MAX_LIGHTING);
+
+  if (*shader == nullptr) {
+    std::string create_info_name = "workbench_composite";
+    create_info_name += workbench_lighting_mode_to_str(light);
+    *shader = GPU_shader_create_from_info_name(create_info_name.c_str());
+  }
+  return *shader;
+}
+
+GPUShader *workbench_shader_merge_infront_get(WORKBENCH_PrivateData *UNUSED(wpd))
+{
+  if (e_data.merge_infront_sh == nullptr) {
+    e_data.merge_infront_sh = GPU_shader_create_from_info_name("workbench_merge_infront");
+  }
+  return e_data.merge_infront_sh;
+}
+
+GPUShader *workbench_shader_transparent_resolve_get(WORKBENCH_PrivateData *UNUSED(wpd))
+{
+  if (e_data.oit_resolve_sh == nullptr) {
+    e_data.oit_resolve_sh = GPU_shader_create_from_info_name("workbench_transparent_resolve");
+  }
+  return e_data.oit_resolve_sh;
+}
+
+static GPUShader *workbench_shader_shadow_pass_get_ex(bool depth_pass, bool manifold, bool cap)
+{
+  struct GPUShader **shader = (depth_pass) ? &e_data.shadow_depth_pass_sh[manifold] :
+                                             &e_data.shadow_depth_fail_sh[manifold][cap];
+
+  if (*shader == nullptr) {
+    std::string create_info_name = "workbench_shadow";
+    create_info_name += (depth_pass) ? "_pass" : "_fail";
+    create_info_name += (manifold) ? "_manifold" : "_no_manifold";
+    create_info_name += (cap) ? "_caps" : "_no_caps";
+#if DEBUG_SHADOW_VOLUME
+    create_info_name += "_debug";
+#endif
+    *shader = GPU_shader_create_from_info_name(create_info_name.c_str());
+  }
+  return *shader;
+}
+
+GPUShader *workbench_shader_shadow_pass_get(bool manifold)
+{
+  return workbench_shader_shadow_pass_get_ex(true, manifold, false);
+}
+
+GPUShader *workbench_shader_shadow_fail_get(bool manifold, bool cap)
+{
+  return workbench_shader_shadow_pass_get_ex(false, manifold, cap);
+}
+
+GPUShader *workbench_shader_cavity_get(bool cavity, bool curvature)
+{
+  BLI_assert(cavity || curvature);
+  struct GPUShader **shader = &e_data.cavity_sh[cavity][curvature];
+
+  if (*shader == nullptr) {
+    std::string create_info_name = "workbench_effect";
+    create_info_name += (cavity) ? "_cavity" : "";
+    create_info_name += (curvature) ? "_curvature" : "";
+    *shader = GPU_shader_create_from_info_name(create_info_name.c_str());
+  }
+  return *shader;
+}
+
+GPUShader *workbench_shader_outline_get(void)
+{
+  if (e_data.outline_sh == nullptr) {
+    e_data.outline_sh = GPU_shader_create_from_info_name("workbench_effect_outline");
+  }
+  return e_data.outline_sh;
+}
+
+void workbench_shader_depth_of_field_get(GPUShader **prepare_sh,
+                                         GPUShader **downsample_sh,
+                                         GPUShader **blur1_sh,
+                                         GPUShader **blur2_sh,
+                                         GPUShader **resolve_sh)
+{
+  if (e_data.dof_prepare_sh == nullptr) {
+    e_data.dof_prepare_sh = GPU_shader_create_from_info_name("workbench_effect_dof_prepare");
+    e_data.dof_downsample_sh = GPU_shader_create_from_info_name("workbench_effect_dof_downsample");
+#if 0 /* TODO(fclem): finish COC min_max optimization */
+    e_data.dof_flatten_v_sh = GPU_shader_create_from_info_name("workbench_effect_dof_flatten_v");
+    e_data.dof_flatten_h_sh = GPU_shader_create_from_info_name("workbench_effect_dof_flatten_h");
+    e_data.dof_dilate_v_sh = GPU_shader_create_from_info_name("workbench_effect_dof_dilate_v");
+    e_data.dof_dilate_h_sh = GPU_shader_create_from_info_name("workbench_effect_dof_dilate_h");
+#endif
+    e_data.dof_blur1_sh = GPU_shader_create_from_info_name("workbench_effect_dof_blur1");
+    e_data.dof_blur2_sh = GPU_shader_create_from_info_name("workbench_effect_dof_blur2");
+    e_data.dof_resolve_sh = GPU_shader_create_from_info_name("workbench_effect_dof_resolve");
+  }
+
+  *prepare_sh = e_data.dof_prepare_sh;
+  *downsample_sh = e_data.dof_downsample_sh;
+  *blur1_sh = e_data.dof_blur1_sh;
+  *blur2_sh = e_data.dof_blur2_sh;
+  *resolve_sh = e_data.dof_resolve_sh;
+}
+
+GPUShader *workbench_shader_antialiasing_accumulation_get(void)
+{
+  if (e_data.aa_accum_sh == nullptr) {
+    e_data.aa_accum_sh = GPU_shader_create_from_info_name("workbench_taa");
+  }
+  return e_data.aa_accum_sh;
+}
+
+GPUShader *workbench_shader_antialiasing_get(int stage)
+{
+  BLI_assert(stage < 3);
+  GPUShader **shader = &e_data.smaa_sh[stage];
+
+  if (*shader == nullptr) {
+    std::string create_info_name = "workbench_smaa_stage_";
+    create_info_name += std::to_string(stage);
+    *shader = GPU_shader_create_from_info_name(create_info_name.c_str());
+  }
+  return e_data.smaa_sh[stage];
+}
+
+GPUShader *workbench_shader_volume_get(bool slice,
+                                       bool coba,
+                                       eWORKBENCH_VolumeInterpType interp_type,
+                                       bool smoke)
+{
+  GPUShader **shader = &e_data.volume_sh[slice][coba][interp_type][smoke];
+
+  if (*shader == nullptr) {
+    std::string create_info_name = "workbench_volume";
+    create_info_name += (smoke) ? "_smoke" : "_object";
+    create_info_name += workbench_volume_interp_to_str(interp_type);
+    create_info_name += (coba) ? "_coba" : "_no_coba";
+    create_info_name += (slice) ? "_slice" : "_no_slice";
+    *shader = GPU_shader_create_from_info_name(create_info_name.c_str());
+  }
+  return *shader;
+}
+
+/** \} */
+
+/* -------------------------------------------------------------------- */
+/** \name Cleanup
+ * \{ */
+
+void workbench_shader_free(void)
+{
+  for (int j = 0; j < sizeof(e_data.opaque_prepass_sh_cache) / sizeof(void *); j++) {
+    struct GPUShader **sh_array = &e_data.opaque_prepass_sh_cache[0][0][0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < sizeof(e_data.transp_prepass_sh_cache) / sizeof(void *); j++) {
+    struct GPUShader **sh_array = &e_data.transp_prepass_sh_cache[0][0][0][0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < ARRAY_SIZE(e_data.opaque_composite_sh); j++) {
+    struct GPUShader **sh_array = &e_data.opaque_composite_sh[0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < ARRAY_SIZE(e_data.shadow_depth_pass_sh); j++) {
+    struct GPUShader **sh_array = &e_data.shadow_depth_pass_sh[0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < sizeof(e_data.shadow_depth_fail_sh) / sizeof(void *); j++) {
+    struct GPUShader **sh_array = &e_data.shadow_depth_fail_sh[0][0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < sizeof(e_data.cavity_sh) / sizeof(void *); j++) {
+    struct GPUShader **sh_array = &e_data.cavity_sh[0][0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < ARRAY_SIZE(e_data.smaa_sh); j++) {
+    struct GPUShader **sh_array = &e_data.smaa_sh[0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+  for (int j = 0; j < sizeof(e_data.volume_sh) / sizeof(void *); j++) {
+    struct GPUShader **sh_array = &e_data.volume_sh[0][0][0][0];
+    DRW_SHADER_FREE_SAFE(sh_array[j]);
+  }
+
+  DRW_SHADER_FREE_SAFE(e_data.oit_resolve_sh);
+  DRW_SHADER_FREE_SAFE(e_data.outline_sh);
+  DRW_SHADER_FREE_SAFE(e_data.merge_infront_sh);
+
+  DRW_SHADER_FREE_SAFE(e_data.dof_prepare_sh);
+  DRW_SHADER_FREE_SAFE(e_data.dof_downsample_sh);
+  DRW_SHADER_FREE_SAFE(e_data.dof_blur1_sh);
+  DRW_SHADER_FREE_SAFE(e_data.dof_blur2_sh);
+  DRW_SHADER_FREE_SAFE(e_data.dof_resolve_sh);
+
+  DRW_SHADER_FREE_SAFE(e_data.aa_accum_sh);
+}
+
+/** \} */

source/blender/draw/engines/workbench/workbench_shader_shared.h

@@ -45,4 +45,3 @@ struct WorldData {
 };
 
 #define viewport_size_inv viewport_size.zw
-#define packed_rough_metal roughness

source/blender/draw/engines/workbench/workbench_transparent.c

@@ -65,7 +65,7 @@ void workbench_transparent_engine_init(WORKBENCH_Data *data)
 static void workbench_transparent_lighting_uniforms(WORKBENCH_PrivateData *wpd,
                                                     DRWShadingGroup *grp)
 {
-  DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+  DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
   DRW_shgroup_uniform_bool_copy(grp, "forceShadowing", false);
 
   if (STUDIOLIGHT_TYPE_MATCAP_ENABLED(wpd)) {
@@ -76,8 +76,8 @@ static void workbench_transparent_lighting_uniforms(WORKBENCH_PrivateData *wpd,
     struct GPUTexture *spec_tx = wpd->studio_light->matcap_specular.gputexture;
     const bool use_spec = workbench_is_specular_highlight_enabled(wpd);
     spec_tx = (use_spec && spec_tx) ? spec_tx : diff_tx;
-    DRW_shgroup_uniform_texture(grp, "matcapDiffuseImage", diff_tx);
-    DRW_shgroup_uniform_texture(grp, "matcapSpecularImage", spec_tx);
+    DRW_shgroup_uniform_texture(grp, "matcap_diffuse_tx", diff_tx);
+    DRW_shgroup_uniform_texture(grp, "matcap_specular_tx", spec_tx);
   }
 }
 
@@ -111,25 +111,25 @@ void workbench_transparent_cache_init(WORKBENCH_Data *vedata)
         sh = workbench_shader_transparent_get(wpd, data);
 
         wpd->prepass[transp][infront][data].common_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", -1);
         workbench_transparent_lighting_uniforms(wpd, grp);
 
         wpd->prepass[transp][infront][data].vcol_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", 0); /* Default material. (uses vcol) */
 
         sh = workbench_shader_transparent_image_get(wpd, data, false);
 
         wpd->prepass[transp][infront][data].image_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", 0); /* Default material. */
         workbench_transparent_lighting_uniforms(wpd, grp);
 
         sh = workbench_shader_transparent_image_get(wpd, data, true);
 
         wpd->prepass[transp][infront][data].image_tiled_shgrp = grp = DRW_shgroup_create(sh, pass);
-        DRW_shgroup_uniform_block(grp, "material_block", wpd->material_ubo_curr);
+        DRW_shgroup_uniform_block(grp, "materials_data", wpd->material_ubo_curr);
         DRW_shgroup_uniform_int_copy(grp, "materialIndex", 0); /* Default material. */
         workbench_transparent_lighting_uniforms(wpd, grp);
       }

source/blender/draw/engines/workbench/workbench_volume.c

@@ -129,7 +129,7 @@ static void workbench_volume_modifier_cache_populate(WORKBENCH_Data *vedata,
     float step_length = max_ff(1e-16f, dim[axis] * 0.05f);
 
     grp = DRW_shgroup_create(sh, vedata->psl->volume_ps);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
     DRW_shgroup_uniform_float_copy(grp, "slicePosition", fds->slice_depth);
     DRW_shgroup_uniform_int_copy(grp, "sliceAxis", axis);
     DRW_shgroup_uniform_float_copy(grp, "stepLength", step_length);
@@ -148,7 +148,7 @@ static void workbench_volume_modifier_cache_populate(WORKBENCH_Data *vedata,
     step_length = len_v3(dim);
 
     grp = DRW_shgroup_create(sh, vedata->psl->volume_ps);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
     DRW_shgroup_uniform_int_copy(grp, "samplesLen", max_slice);
     DRW_shgroup_uniform_float_copy(grp, "stepLength", step_length);
     DRW_shgroup_uniform_float_copy(grp, "noiseOfs", noise_ofs);
@@ -272,7 +272,7 @@ static void workbench_volume_object_cache_populate(WORKBENCH_Data *vedata,
     const float slice_position = volume->display.slice_depth;
 
     grp = DRW_shgroup_create(sh, vedata->psl->volume_ps);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
     DRW_shgroup_uniform_float_copy(grp, "slicePosition", slice_position);
     DRW_shgroup_uniform_int_copy(grp, "sliceAxis", axis);
     DRW_shgroup_uniform_float_copy(grp, "stepLength", step_length);
@@ -299,7 +299,7 @@ static void workbench_volume_object_cache_populate(WORKBENCH_Data *vedata,
 
     /* Set uniforms. */
     grp = DRW_shgroup_create(sh, vedata->psl->volume_ps);
-    DRW_shgroup_uniform_block(grp, "world_block", wpd->world_ubo);
+    DRW_shgroup_uniform_block(grp, "world_data", wpd->world_ubo);
     DRW_shgroup_uniform_int_copy(grp, "samplesLen", max_slice);
     DRW_shgroup_uniform_float_copy(grp, "stepLength", step_length);
     DRW_shgroup_uniform_float_copy(grp, "noiseOfs", noise_ofs);

source/blender/draw/intern/draw_manager_data.c

@@ -1248,6 +1248,17 @@ static void drw_shgroup_init(DRWShadingGroup *shgroup, GPUShader *shader)
   int chunkid_location = GPU_shader_get_builtin_uniform(shader, GPU_UNIFORM_RESOURCE_CHUNK);
   int resourceid_location = GPU_shader_get_builtin_uniform(shader, GPU_UNIFORM_RESOURCE_ID);
 
+  /* TODO(fclem) Will take the place of the above after the GPUShaderCreateInfo port. */
+  if (view_ubo_location == -1) {
+    view_ubo_location = GPU_shader_get_builtin_block(shader, GPU_UNIFORM_BLOCK_DRW_VIEW);
+  }
+  if (model_ubo_location == -1) {
+    model_ubo_location = GPU_shader_get_builtin_block(shader, GPU_UNIFORM_BLOCK_DRW_MODEL);
+  }
+  if (info_ubo_location == -1) {
+    info_ubo_location = GPU_shader_get_builtin_block(shader, GPU_UNIFORM_BLOCK_DRW_INFOS);
+  }
+
   if (chunkid_location != -1) {
     drw_shgroup_uniform_create_ex(
         shgroup, chunkid_location, DRW_UNIFORM_RESOURCE_CHUNK, NULL, 0, 0, 1);

source/blender/draw/intern/draw_shader_shared.h

@@ -16,7 +16,7 @@ struct ViewInfos {
   float4x4 winmat;
   float4x4 wininv;
 
-  float4 clipplanes[6];
+  float4 clip_planes[6];
   float4 viewvecs[2];
   /* Should not be here. Not view dependent (only main view). */
   float4 viewcamtexcofac;
@@ -30,7 +30,7 @@ BLI_STATIC_ASSERT_ALIGN(ViewInfos, 16)
 #define ViewMatrixInverse drw_view.viewinv
 #define ProjectionMatrix drw_view.winmat
 #define ProjectionMatrixInverse drw_view.wininv
-#define clipPlanes drw_view.clipplanes
+#define clipPlanes drw_view.clip_planes
 #define ViewVecs drw_view.viewvecs
 #define CameraTexCoFactors drw_view.viewcamtexcofac
 
@@ -39,3 +39,14 @@ struct ObjectMatrices {
   float4x4 drw_modelMatrixInverse;
 };
 BLI_STATIC_ASSERT_ALIGN(ViewInfos, 16)
+
+struct ObjectInfos {
+  float4 drw_OrcoTexCoFactors[2];
+  float4 drw_ObjectColor;
+  float4 drw_Infos;
+};
+BLI_STATIC_ASSERT_ALIGN(ViewInfos, 16)
+
+#define OrcoTexCoFactors (drw_infos[resource_id].drw_OrcoTexCoFactors)
+#define ObjectInfo (drw_infos[resource_id].drw_Infos)
+#define ObjectColor (drw_infos[resource_id].drw_ObjectColor)