Merge branch 'blender-v3.3-release'

11 files changed, 157 insertions, 78 deletions

diff --git a/intern/cycles/blender/curves.cpp b/intern/cycles/blender/curves.cpp
index 04ba7c53878..59e630eef63 100644
--- a/intern/cycles/blender/curves.cpp
+++ b/intern/cycles/blender/curves.cpp
@@ -707,6 +707,21 @@ static void attr_create_motion(Hair *hair, BL::Attribute &b_attribute, const flo
   }
 }
 
+static void attr_create_uv(AttributeSet &attributes,
+                           BL::Curves &b_curves,
+                           BL::Attribute &b_attribute,
+                           const ustring name)
+{
+  BL::Float2Attribute b_float2_attribute{b_attribute};
+  Attribute *attr = attributes.add(ATTR_STD_UV, name);
+
+  float2 *data = attr->data_float2();
+  fill_generic_attribute(b_curves, data, ATTR_ELEMENT_CURVE, [&](int i) {
+    BL::Array<float, 2> v = b_float2_attribute.data[i].vector();
+    return make_float2(v[0], v[1]);
+  });
+}
+
 static void attr_create_generic(Scene *scene,
                                 Hair *hair,
                                 BL::Curves &b_curves,
@@ -715,12 +730,26 @@ static void attr_create_generic(Scene *scene,
 {
   AttributeSet &attributes = hair->attributes;
   static const ustring u_velocity("velocity");
+  const bool need_uv = hair->need_attribute(scene, ATTR_STD_UV);
+  bool have_uv = false;
 
   for (BL::Attribute &b_attribute : b_curves.attributes) {
     const ustring name{b_attribute.name().c_str()};
 
+    const BL::Attribute::domain_enum b_domain = b_attribute.domain();
+    const BL::Attribute::data_type_enum b_data_type = b_attribute.data_type();
+
     if (need_motion && name == u_velocity) {
       attr_create_motion(hair, b_attribute, motion_scale);
+      continue;
+    }
+
+    /* Weak, use first float2 attribute as standard UV. */
+    if (need_uv && !have_uv && b_data_type == BL::Attribute::data_type_FLOAT2 &&
+        b_domain == BL::Attribute::domain_CURVE) {
+      attr_create_uv(attributes, b_curves, b_attribute, name);
+      have_uv = true;
+      continue;
     }
 
     if (!hair->need_attribute(scene, name)) {
@@ -730,9 +759,6 @@ static void attr_create_generic(Scene *scene,
       continue;
     }
 
-    const BL::Attribute::domain_enum b_domain = b_attribute.domain();
-    const BL::Attribute::data_type_enum b_data_type = b_attribute.data_type();
-
     AttributeElement element = ATTR_ELEMENT_NONE;
     switch (b_domain) {
       case BL::Attribute::domain_POINT:
@@ -844,79 +870,84 @@ static void export_hair_curves(Scene *scene,
 {
   /* TODO: optimize so we can straight memcpy arrays from Blender? */
 
+  const int num_keys = b_curves.points.length();
+  const int num_curves = b_curves.curves.length();
+
+  hair->resize_curves(num_curves, num_keys);
+
+  float3 *curve_keys = hair->get_curve_keys().data();
+  float *curve_radius = hair->get_curve_radius().data();
+  int *curve_first_key = hair->get_curve_first_key().data();
+  int *curve_shader = hair->get_curve_shader().data();
+
   /* Add requested attributes. */
-  Attribute *attr_intercept = NULL;
-  Attribute *attr_length = NULL;
-  Attribute *attr_random = NULL;
+  float *attr_intercept = NULL;
+  float *attr_length = NULL;
+  float *attr_random = NULL;
 
   if (hair->need_attribute(scene, ATTR_STD_CURVE_INTERCEPT)) {
-    attr_intercept = hair->attributes.add(ATTR_STD_CURVE_INTERCEPT);
+    attr_intercept = hair->attributes.add(ATTR_STD_CURVE_INTERCEPT)->data_float();
   }
   if (hair->need_attribute(scene, ATTR_STD_CURVE_LENGTH)) {
-    attr_length = hair->attributes.add(ATTR_STD_CURVE_LENGTH);
+    attr_length = hair->attributes.add(ATTR_STD_CURVE_LENGTH)->data_float();
   }
   if (hair->need_attribute(scene, ATTR_STD_CURVE_RANDOM)) {
-    attr_random = hair->attributes.add(ATTR_STD_CURVE_RANDOM);
+    attr_random = hair->attributes.add(ATTR_STD_CURVE_RANDOM)->data_float();
   }
 
-  /* Reserve memory. */
-  const int num_keys = b_curves.points.length();
-  const int num_curves = b_curves.curves.length();
-
-  hair->reserve_curves(num_curves, num_keys);
-
   BL::FloatVectorAttribute b_attr_position = find_curves_position_attribute(b_curves);
   std::optional<BL::FloatAttribute> b_attr_radius = find_curves_radius_attribute(b_curves);
 
   /* Export curves and points. */
-  vector<float> points_length;
-
   for (int i = 0; i < num_curves; i++) {
     const int first_point_index = b_curves.curve_offset_data[i].value();
     const int num_points = b_curves.curve_offset_data[i + 1].value() - first_point_index;
 
     float3 prev_co = zero_float3();
     float length = 0.0f;
-    if (attr_intercept) {
-      points_length.clear();
-      points_length.reserve(num_points);
-    }
 
     /* Position and radius. */
-    for (int i = 0; i < num_points; i++) {
-      const float3 co = get_float3(b_attr_position.data[first_point_index + i].vector());
-      const float radius = b_attr_radius ? b_attr_radius->data[first_point_index + i].value() :
-                                           0.005f;
-      hair->add_curve_key(co, radius);
-
-      if (attr_intercept) {
-        if (i > 0) {
+    for (int j = 0; j < num_points; j++) {
+      const int point_offset = first_point_index + j;
+      const float3 co = get_float3(b_attr_position.data[point_offset].vector());
+      const float radius = b_attr_radius ? b_attr_radius->data[point_offset].value() : 0.0f;
+
+      curve_keys[point_offset] = co;
+      curve_radius[point_offset] = radius;
+
+      if (attr_length || attr_intercept) {
+        if (j > 0) {
           length += len(co - prev_co);
-          points_length.push_back(length);
         }
         prev_co = co;
+
+        if (attr_intercept) {
+          attr_intercept[point_offset] = length;
+        }
       }
     }
 
     /* Normalized 0..1 attribute along curve. */
-    if (attr_intercept) {
-      for (int i = 0; i < num_points; i++) {
-        attr_intercept->add((length == 0.0f) ? 0.0f : points_length[i] / length);
+    if (attr_intercept && length > 0.0f) {
+      for (int j = 1; j < num_points; j++) {
+        const int point_offset = first_point_index + j;
+        attr_intercept[point_offset] /= length;
       }
     }
 
+    /* Curve length. */
     if (attr_length) {
-      attr_length->add(length);
+      attr_length[i] = length;
     }
 
     /* Random number per curve. */
     if (attr_random != NULL) {
-      attr_random->add(hash_uint2_to_float(i, 0));
+      attr_random[i] = hash_uint2_to_float(i, 0);
     }
 
     /* Curve. */
-    const int shader_index = 0;
-    hair->add_curve(first_point_index, shader_index);
+    curve_shader[i] = 0;
+    curve_first_key[i] = first_point_index;
   }
 
   attr_create_generic(scene, hair, b_curves, need_motion, motion_scale);
diff --git a/intern/cycles/blender/display_driver.cpp b/intern/cycles/blender/display_driver.cpp
index a1bc064be68..61cd88fb433 100644
--- a/intern/cycles/blender/display_driver.cpp
+++ b/intern/cycles/blender/display_driver.cpp
@@ -913,8 +913,6 @@ void BlenderDisplayDriver::flush()
 void BlenderDisplayDriver::draw(const Params &params)
 {
   /* See do_update_begin() for why no locking is required here. */
-  const bool transparent = true;  // TODO(sergey): Derive this from Film.
-
   if (use_gl_context_) {
     gl_context_mutex_.lock();
   }
@@ -935,10 +933,8 @@ void BlenderDisplayDriver::draw(const Params &params)
     glWaitSync((GLsync)gl_upload_sync_, 0, GL_TIMEOUT_IGNORED);
   }
 
-  if (transparent) {
-    glEnable(GL_BLEND);
-    glBlendFunc(GL_ONE, GL_ONE_MINUS_SRC_ALPHA);
-  }
+  glEnable(GL_BLEND);
+  glBlendFunc(GL_ONE, GL_ONE_MINUS_SRC_ALPHA);
 
   glActiveTexture(GL_TEXTURE0);
 
@@ -975,9 +971,7 @@ void BlenderDisplayDriver::draw(const Params &params)
 
   glDeleteVertexArrays(1, &vertex_array_object);
 
-  if (transparent) {
-    glDisable(GL_BLEND);
-  }
+  glDisable(GL_BLEND);
 
   gl_render_sync_ = glFenceSync(GL_SYNC_GPU_COMMANDS_COMPLETE, 0);
   glFlush();
diff --git a/intern/cycles/bvh/bvh.h b/intern/cycles/bvh/bvh.h
index be390f8a673..19ebf7f68ba 100644
--- a/intern/cycles/bvh/bvh.h
+++ b/intern/cycles/bvh/bvh.h
@@ -74,6 +74,13 @@ class BVH {
   {
   }
 
+  virtual void replace_geometry(const vector<Geometry *> &geometry,
+                                const vector<Object *> &objects)
+  {
+    this->geometry = geometry;
+    this->objects = objects;
+  }
+
  protected:
   BVH(const BVHParams &params,
       const vector<Geometry *> &geometry,
diff --git a/intern/cycles/bvh/multi.cpp b/intern/cycles/bvh/multi.cpp
index 7211720b56b..d9ee2fce966 100644
--- a/intern/cycles/bvh/multi.cpp
+++ b/intern/cycles/bvh/multi.cpp
@@ -21,4 +21,12 @@ BVHMulti::~BVHMulti()
   }
 }
 
+void BVHMulti::replace_geometry(const vector<Geometry *> &geometry,
+                                const vector<Object *> &objects)
+{
+  foreach (BVH *bvh, sub_bvhs) {
+    bvh->replace_geometry(geometry, objects);
+  }
+}
+
 CCL_NAMESPACE_END
diff --git a/intern/cycles/bvh/multi.h b/intern/cycles/bvh/multi.h
index 824899f3101..aafbfae19e0 100644
--- a/intern/cycles/bvh/multi.h
+++ b/intern/cycles/bvh/multi.h
@@ -19,6 +19,9 @@ class BVHMulti : public BVH {
            const vector<Geometry *> &geometry,
            const vector<Object *> &objects);
   virtual ~BVHMulti();
+
+  virtual void replace_geometry(const vector<Geometry *> &geometry,
+                                const vector<Object *> &objects);
 };
 
 CCL_NAMESPACE_END
diff --git a/intern/cycles/kernel/device/cpu/bvh.h b/intern/cycles/kernel/device/cpu/bvh.h
index 6c06232a692..d9267e1cd6d 100644
--- a/intern/cycles/kernel/device/cpu/bvh.h
+++ b/intern/cycles/kernel/device/cpu/bvh.h
@@ -114,27 +114,37 @@ ccl_device_inline bool kernel_embree_is_self_intersection(const KernelGlobals kg
                                                           const RTCHit *hit,
                                                           const Ray *ray)
 {
-  bool status = false;
+  int object, prim;
+
   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)) {
+    object = hit->instID[0] / 2;
+    if ((ray->self.object == object) || (ray->self.light_object == object)) {
       RTCScene inst_scene = (RTCScene)rtcGetGeometryUserData(
           rtcGetGeometry(kernel_data.device_bvh, 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);
+      prim = hit->primID +
+             (intptr_t)rtcGetGeometryUserData(rtcGetGeometry(inst_scene, hit->geomID));
+    }
+    else {
+      return false;
     }
   }
   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.device_bvh, hit->geomID));
-      status = intersection_skip_self_shadow(ray->self, oID, pID);
+    object = hit->geomID / 2;
+    if ((ray->self.object == object) || (ray->self.light_object == object)) {
+      prim = hit->primID +
+             (intptr_t)rtcGetGeometryUserData(rtcGetGeometry(kernel_data.device_bvh, hit->geomID));
     }
+    else {
+      return false;
+    }
+  }
+
+  const bool is_hair = hit->geomID & 1;
+  if (is_hair) {
+    prim = kernel_data_fetch(curve_segments, prim).prim;
   }
 
-  return status;
+  return intersection_skip_self_shadow(ray->self, object, prim);
 }
 
 ccl_device_inline void kernel_embree_convert_hit(KernelGlobals kg,
diff --git a/intern/cycles/kernel/device/metal/kernel.metal b/intern/cycles/kernel/device/metal/kernel.metal
index 3d173b0d601..3de8c069c30 100644
--- a/intern/cycles/kernel/device/metal/kernel.metal
+++ b/intern/cycles/kernel/device/metal/kernel.metal
@@ -180,11 +180,6 @@ bool metalrt_shadow_all_hit(constant KernelParamsMetal &launch_params_metal,
   }
 #  endif
 
-  if (intersection_skip_self_shadow(payload.self, object, prim)) {
-    /* continue search */
-    return true;
-  }
-
   const float u = barycentrics.x;
   const float v = barycentrics.y;
   int type = 0;
@@ -205,6 +200,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;
+  }
+
 #  ifndef __TRANSPARENT_SHADOWS__
   /* No transparent shadows support compiled in, make opaque. */
   payload.result = true;
@@ -346,6 +346,14 @@ inline TReturnType metalrt_visibility_test(
   }
 #endif
 
+  if (intersection_type == METALRT_HIT_TRIANGLE) {
+  }
+#  ifdef __HAIR__
+  else {
+    prim = kernel_data_fetch(curve_segments, prim).prim;
+  }
+#  endif
+
   /* Shadow ray early termination. */
   if (visibility & PATH_RAY_SHADOW_OPAQUE) {
     if (intersection_skip_self_shadow(payload.self, object, prim)) {
diff --git a/intern/cycles/kernel/device/optix/bvh.h b/intern/cycles/kernel/device/optix/bvh.h
index d1d342f6034..fb9907709ce 100644
--- a/intern/cycles/kernel/device/optix/bvh.h
+++ b/intern/cycles/kernel/device/optix/bvh.h
@@ -143,14 +143,10 @@ 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()) {
+    /* Triangle. */
     const float2 barycentrics = optixGetTriangleBarycentrics();
     u = barycentrics.x;
     v = barycentrics.y;
@@ -158,6 +154,7 @@ extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit()
   }
 #  ifdef __HAIR__
   else if ((optixGetHitKind() & (~PRIMITIVE_MOTION)) != PRIMITIVE_POINT) {
+    /* Curve. */
     u = __uint_as_float(optixGetAttribute_0());
     v = __uint_as_float(optixGetAttribute_1());
 
@@ -174,11 +171,17 @@ extern "C" __global__ void __anyhit__kernel_optix_shadow_all_hit()
   }
 #  endif
   else {
+    /* Point. */
     type = kernel_data_fetch(objects, object).primitive_type;
     u = 0.0f;
     v = 0.0f;
   }
 
+  ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
+  if (intersection_skip_self_shadow(ray->self, object, prim)) {
+    return optixIgnoreIntersection();
+  }
+
 #  ifndef __TRANSPARENT_SHADOWS__
   /* No transparent shadows support compiled in, make opaque. */
   optixSetPayload_5(true);
@@ -307,7 +310,17 @@ extern "C" __global__ void __anyhit__kernel_optix_visibility_test()
   }
 #endif
 
-  const int prim = optixGetPrimitiveIndex();
+  int prim = optixGetPrimitiveIndex();
+  if (optixIsTriangleHit()) {
+    /* Triangle. */
+  }
+#ifdef __HAIR__
+  else if ((optixGetHitKind() & (~PRIMITIVE_MOTION)) != PRIMITIVE_POINT) {
+    /* Curve. */
+    prim = kernel_data_fetch(curve_segments, prim).prim;
+  }
+#endif
+
   ccl_private Ray *const ray = get_payload_ptr_6<Ray>();
 
   if (visibility & PATH_RAY_SHADOW_OPAQUE) {
diff --git a/intern/cycles/kernel/integrator/shade_surface.h b/intern/cycles/kernel/integrator/shade_surface.h
index 1514b3956ad..70b20a93b6a 100644
--- a/intern/cycles/kernel/integrator/shade_surface.h
+++ b/intern/cycles/kernel/integrator/shade_surface.h
@@ -43,11 +43,9 @@ ccl_device_forceinline bool integrate_surface_holdout(KernelGlobals kg,
   if (((sd->flag & SD_HOLDOUT) || (sd->object_flag & SD_OBJECT_HOLDOUT_MASK)) &&
       (path_flag & PATH_RAY_TRANSPARENT_BACKGROUND)) {
     const float3 holdout_weight = shader_holdout_apply(kg, sd);
-    if (kernel_data.background.transparent) {
-      const float3 throughput = INTEGRATOR_STATE(state, path, throughput);
-      const float transparent = average(holdout_weight * throughput);
-      kernel_accum_holdout(kg, state, path_flag, transparent, render_buffer);
-    }
+    const float3 throughput = INTEGRATOR_STATE(state, path, throughput);
+    const float transparent = average(holdout_weight * throughput);
+    kernel_accum_holdout(kg, state, path_flag, transparent, render_buffer);
     if (isequal(holdout_weight, one_float3())) {
       return false;
     }
diff --git a/intern/cycles/scene/geometry.cpp b/intern/cycles/scene/geometry.cpp
index 67ff118692e..ae8dcaa43b6 100644
--- a/intern/cycles/scene/geometry.cpp
+++ b/intern/cycles/scene/geometry.cpp
@@ -217,8 +217,7 @@ void Geometry::compute_bvh(Device *device,
     if (bvh && !need_update_rebuild) {
       progress->set_status(msg, "Refitting BVH");
 
-      bvh->geometry = geometry;
-      bvh->objects = objects;
+      bvh->replace_geometry(geometry, objects);
 
       device->build_bvh(bvh, *progress, true);
     }
diff --git a/source/blender/render/intern/pipeline.c b/source/blender/render/intern/pipeline.c
index 1c42467bc3d..30e8cfa5c17 100644
--- a/source/blender/render/intern/pipeline.c
+++ b/source/blender/render/intern/pipeline.c
@@ -199,14 +199,20 @@ static void stats_background(void *UNUSED(arg), RenderStats *rs)
   megs_used_memory = (mem_in_use) / (1024.0 * 1024.0);
   megs_peak_memory = (peak_memory) / (1024.0 * 1024.0);
 
+  BLI_timecode_string_from_time_simple(
+      info_time_str, sizeof(info_time_str), PIL_check_seconds_timer() - rs->starttime);
+
+  /* Compositor calls this from multiple threads, mutex lock to ensure we don't
+   * get garbled output. */
+  static ThreadMutex mutex = BLI_MUTEX_INITIALIZER;
+  BLI_mutex_lock(&mutex);
+
   fprintf(stdout,
           TIP_("Fra:%d Mem:%.2fM (Peak %.2fM) "),
           rs->cfra,
           megs_used_memory,
           megs_peak_memory);
 
-  BLI_timecode_string_from_time_simple(
-      info_time_str, sizeof(info_time_str), PIL_check_seconds_timer() - rs->starttime);
   fprintf(stdout, TIP_("| Time:%s | "), info_time_str);
 
   fprintf(stdout, "%s", rs->infostr);
@@ -220,6 +226,8 @@ static void stats_background(void *UNUSED(arg), RenderStats *rs)
 
   fputc('\n', stdout);
   fflush(stdout);
+
+  BLI_mutex_unlock(&mutex);
 }
 
 void RE_FreeRenderResult(RenderResult *rr)