Welcome to mirror list, hosted at ThFree Co, Russian Federation.

git.blender.org/blender.git - Unnamed repository; edit this file 'description' to name the repository.
summaryrefslogtreecommitdiff
path: root/intern/cycles
diff options
context:
space:
mode:
Diffstat (limited to 'intern/cycles')
-rw-r--r--intern/cycles/blender/blender_mesh.cpp46
-rw-r--r--intern/cycles/blender/blender_session.cpp3
-rw-r--r--intern/cycles/device/CMakeLists.txt1
-rw-r--r--intern/cycles/device/opencl/device_opencl_impl.cpp8
-rw-r--r--intern/cycles/kernel/kernel_film.h2
-rw-r--r--intern/cycles/kernel/kernels/opencl/kernel_bake.cl14
-rw-r--r--intern/cycles/kernel/osl/osl_globals.h1
-rw-r--r--intern/cycles/kernel/osl/osl_services.cpp8
-rw-r--r--intern/cycles/render/buffers.cpp15
-rw-r--r--intern/cycles/render/buffers.h2
-rw-r--r--intern/cycles/render/film.cpp2
11 files changed, 57 insertions, 45 deletions
diff --git a/intern/cycles/blender/blender_mesh.cpp b/intern/cycles/blender/blender_mesh.cpp
index f4354d5166e..e40e1f5f001 100644
--- a/intern/cycles/blender/blender_mesh.cpp
+++ b/intern/cycles/blender/blender_mesh.cpp
@@ -923,48 +923,34 @@ static void create_subd_mesh(Scene *scene,
/* Sync */
-static BL::MeshSequenceCacheModifier object_mesh_cache_find(BL::Object &b_ob,
- BL::Scene /*b_scene*/)
+static BL::MeshSequenceCacheModifier object_mesh_cache_find(BL::Object &b_ob)
{
- BL::Object::modifiers_iterator b_mod;
+ if (b_ob.modifiers.length() > 0) {
+ BL::Modifier b_mod = b_ob.modifiers[b_ob.modifiers.length() - 1];
- for (b_ob.modifiers.begin(b_mod); b_mod != b_ob.modifiers.end(); ++b_mod) {
- if (!b_mod->is_a(&RNA_MeshSequenceCacheModifier)) {
- continue;
- }
-
- BL::MeshSequenceCacheModifier mesh_cache = BL::MeshSequenceCacheModifier(*b_mod);
+ if (b_mod.type() == BL::Modifier::type_MESH_SEQUENCE_CACHE) {
+ BL::MeshSequenceCacheModifier mesh_cache = BL::MeshSequenceCacheModifier(b_mod);
- if (MeshSequenceCacheModifier_has_velocity_get(&mesh_cache.ptr)) {
- return mesh_cache;
+ if (MeshSequenceCacheModifier_has_velocity_get(&mesh_cache.ptr)) {
+ return mesh_cache;
+ }
}
}
return BL::MeshSequenceCacheModifier(PointerRNA_NULL);
}
-static void sync_mesh_cached_velocities(BL::Object &b_ob,
- BL::Scene b_scene,
- Scene *scene,
- Mesh *mesh)
+static void sync_mesh_cached_velocities(BL::Object &b_ob, Scene *scene, Mesh *mesh)
{
if (scene->need_motion() == Scene::MOTION_NONE)
return;
- BL::MeshSequenceCacheModifier b_mesh_cache = object_mesh_cache_find(b_ob, b_scene);
+ BL::MeshSequenceCacheModifier b_mesh_cache = object_mesh_cache_find(b_ob);
if (!b_mesh_cache) {
return;
}
- /* Find or add attribute */
- float3 *P = &mesh->verts[0];
- Attribute *attr_mP = mesh->attributes.find(ATTR_STD_MOTION_VERTEX_POSITION);
-
- if (!attr_mP) {
- attr_mP = mesh->attributes.add(ATTR_STD_MOTION_VERTEX_POSITION);
- }
-
if (!MeshSequenceCacheModifier_read_velocity_get(&b_mesh_cache.ptr)) {
return;
}
@@ -975,6 +961,14 @@ static void sync_mesh_cached_velocities(BL::Object &b_ob,
return;
}
+ /* Find or add attribute */
+ float3 *P = &mesh->verts[0];
+ Attribute *attr_mP = mesh->attributes.find(ATTR_STD_MOTION_VERTEX_POSITION);
+
+ if (!attr_mP) {
+ attr_mP = mesh->attributes.add(ATTR_STD_MOTION_VERTEX_POSITION);
+ }
+
/* Only export previous and next frame, we don't have any in between data. */
float motion_times[2] = {-1.0f, 1.0f};
for (int step = 0; step < 2; step++) {
@@ -1071,7 +1065,7 @@ void BlenderSync::sync_mesh(BL::Depsgraph b_depsgraph,
}
/* cached velocities (e.g. from alembic archive) */
- sync_mesh_cached_velocities(b_ob, b_depsgraph.scene(), scene, mesh);
+ sync_mesh_cached_velocities(b_ob, scene, mesh);
/* mesh fluid motion mantaflow */
sync_mesh_fluid_motion(b_ob, scene, mesh);
@@ -1095,7 +1089,7 @@ void BlenderSync::sync_mesh_motion(BL::Depsgraph b_depsgraph,
}
/* Cached motion blur already exported. */
- BL::MeshSequenceCacheModifier mesh_cache = object_mesh_cache_find(b_ob, b_scene);
+ BL::MeshSequenceCacheModifier mesh_cache = object_mesh_cache_find(b_ob);
if (mesh_cache) {
return;
}
diff --git a/intern/cycles/blender/blender_session.cpp b/intern/cycles/blender/blender_session.cpp
index fb704b2a24a..bf9fc784d79 100644
--- a/intern/cycles/blender/blender_session.cpp
+++ b/intern/cycles/blender/blender_session.cpp
@@ -363,7 +363,8 @@ void BlenderSession::do_write_update_render_tile(RenderTile &rtile,
PassType pass_type = BlenderSync::get_pass_type(b_pass);
int components = b_pass.channels();
- rtile.buffers->set_pass_rect(pass_type, components, (float *)b_pass.rect());
+ rtile.buffers->set_pass_rect(
+ pass_type, components, (float *)b_pass.rect(), rtile.num_samples);
}
end_render_result(b_engine, b_rr, false, false, false);
diff --git a/intern/cycles/device/CMakeLists.txt b/intern/cycles/device/CMakeLists.txt
index ca366722eb7..466872e0557 100644
--- a/intern/cycles/device/CMakeLists.txt
+++ b/intern/cycles/device/CMakeLists.txt
@@ -67,6 +67,7 @@ set(LIB
cycles_render
cycles_kernel
cycles_util
+ ${BLENDER_GL_LIBRARIES}
)
if(WITH_CUDA_DYNLOAD)
diff --git a/intern/cycles/device/opencl/device_opencl_impl.cpp b/intern/cycles/device/opencl/device_opencl_impl.cpp
index e851749949d..f0683d12f1f 100644
--- a/intern/cycles/device/opencl/device_opencl_impl.cpp
+++ b/intern/cycles/device/opencl/device_opencl_impl.cpp
@@ -864,6 +864,11 @@ void OpenCLDevice::load_preview_kernels()
bool OpenCLDevice::wait_for_availability(const DeviceRequestedFeatures &requested_features)
{
+ if (requested_features.use_baking) {
+ /* For baking, kernels have already been loaded in load_required_kernels(). */
+ return true;
+ }
+
if (background) {
load_kernel_task_pool.wait_work();
use_preview_kernels = false;
@@ -1933,13 +1938,12 @@ void OpenCLDevice::bake(DeviceTask &task, RenderTile &rtile)
kernel_set_args(kernel, start_arg_index, sample);
enqueue_kernel(kernel, d_w, d_h);
+ clFinish(cqCommandQueue);
rtile.sample = sample + 1;
task.update_progress(&rtile, rtile.w * rtile.h);
}
-
- clFinish(cqCommandQueue);
}
static bool kernel_build_opencl_2(cl_device_id cdDevice)
diff --git a/intern/cycles/kernel/kernel_film.h b/intern/cycles/kernel/kernel_film.h
index 8344f4b4f47..17b69b6198b 100644
--- a/intern/cycles/kernel/kernel_film.h
+++ b/intern/cycles/kernel/kernel_film.h
@@ -47,7 +47,7 @@ ccl_device float4 film_get_pass_result(KernelGlobals *kg,
if (kernel_data.film.use_display_exposure) {
float exposure = kernel_data.film.exposure;
- pass_result *= make_float4(exposure, exposure, exposure, alpha);
+ pass_result *= make_float4(exposure, exposure, exposure, 1.0f);
}
}
else if (display_pass_components == 1) {
diff --git a/intern/cycles/kernel/kernels/opencl/kernel_bake.cl b/intern/cycles/kernel/kernels/opencl/kernel_bake.cl
index 041312b53cb..7b81e387467 100644
--- a/intern/cycles/kernel/kernels/opencl/kernel_bake.cl
+++ b/intern/cycles/kernel/kernels/opencl/kernel_bake.cl
@@ -12,12 +12,11 @@
__kernel void kernel_ocl_bake(
ccl_constant KernelData *data,
- ccl_global uint4 *input,
- ccl_global float4 *output,
+ ccl_global float *buffer,
KERNEL_BUFFER_PARAMS,
- int type, int filter, int sx, int sw, int offset, int sample)
+ int sx, int sy, int sw, int sh, int offset, int stride, int sample)
{
KernelGlobals kglobals, *kg = &kglobals;
@@ -27,12 +26,11 @@ __kernel void kernel_ocl_bake(
kernel_set_buffer_info(kg);
int x = sx + ccl_global_id(0);
+ int y = sy + ccl_global_id(1);
- if(x < sx + sw) {
-#ifdef __NO_BAKING__
- output[x] = make_float4(0.0f, 0.0f, 0.0f, 0.0f);
-#else
- kernel_bake_evaluate(kg, input, output, (ShaderEvalType)type, filter, x, offset, sample);
+ if(x < sx + sw && y < sy + sh) {
+#ifndef __NO_BAKING__
+ kernel_bake_evaluate(kg, buffer, sample, x, y, offset, stride);
#endif
}
}
diff --git a/intern/cycles/kernel/osl/osl_globals.h b/intern/cycles/kernel/osl/osl_globals.h
index c06c9abd4c1..caca3c28c8d 100644
--- a/intern/cycles/kernel/osl/osl_globals.h
+++ b/intern/cycles/kernel/osl/osl_globals.h
@@ -90,6 +90,7 @@ struct OSLTraceData {
ShaderData sd;
bool setup;
bool init;
+ bool hit;
};
/* thread key for thread specific data lookup */
diff --git a/intern/cycles/kernel/osl/osl_services.cpp b/intern/cycles/kernel/osl/osl_services.cpp
index 5292b5f8055..aee1e3a244e 100644
--- a/intern/cycles/kernel/osl/osl_services.cpp
+++ b/intern/cycles/kernel/osl/osl_services.cpp
@@ -1481,6 +1481,7 @@ bool OSLRenderServices::trace(TraceOpt &options,
tracedata->ray = ray;
tracedata->setup = false;
tracedata->init = true;
+ tracedata->hit = false;
tracedata->sd.osl_globals = sd->osl_globals;
KernelGlobals *kg = sd->osl_globals;
@@ -1492,7 +1493,8 @@ bool OSLRenderServices::trace(TraceOpt &options,
/* Raytrace, leaving out shadow opaque to avoid early exit. */
uint visibility = PATH_RAY_ALL_VISIBILITY - PATH_RAY_SHADOW_OPAQUE;
- return scene_intersect(kg, &ray, visibility, &tracedata->isect);
+ tracedata->hit = scene_intersect(kg, &ray, visibility, &tracedata->isect);
+ return tracedata->hit;
}
bool OSLRenderServices::getmessage(OSL::ShaderGlobals *sg,
@@ -1506,9 +1508,9 @@ bool OSLRenderServices::getmessage(OSL::ShaderGlobals *sg,
if (source == u_trace && tracedata->init) {
if (name == u_hit) {
- return set_attribute_int((tracedata->isect.prim != PRIM_NONE), type, derivatives, val);
+ return set_attribute_int(tracedata->hit, type, derivatives, val);
}
- else if (tracedata->isect.prim != PRIM_NONE) {
+ else if (tracedata->hit) {
if (name == u_hitdist) {
float f[3] = {tracedata->isect.t, 0.0f, 0.0f};
return set_attribute_float(f, type, derivatives, val);
diff --git a/intern/cycles/render/buffers.cpp b/intern/cycles/render/buffers.cpp
index b26366af852..3607300cee6 100644
--- a/intern/cycles/render/buffers.cpp
+++ b/intern/cycles/render/buffers.cpp
@@ -459,7 +459,7 @@ bool RenderBuffers::get_pass_rect(
return false;
}
-bool RenderBuffers::set_pass_rect(PassType type, int components, float *pixels)
+bool RenderBuffers::set_pass_rect(PassType type, int components, float *pixels, int samples)
{
if (buffer.data() == NULL) {
return false;
@@ -482,8 +482,17 @@ bool RenderBuffers::set_pass_rect(PassType type, int components, float *pixels)
assert(pass.components == components);
for (int i = 0; i < size; i++, out += pass_stride, pixels += components) {
- for (int j = 0; j < components; j++) {
- out[j] = pixels[j];
+ if (pass.filter) {
+ /* Scale by the number of samples, inverse of what we do in get_pass_rect.
+ * A better solution would be to remove the need for set_pass_rect entirely,
+ * and change baking to bake multiple objects in a tile at once. */
+ for (int j = 0; j < components; j++) {
+ out[j] = pixels[j] * samples;
+ }
+ }
+ else {
+ /* For non-filtered passes just straight copy, these may contain non-float data. */
+ memcpy(out, pixels, sizeof(float) * components);
}
}
diff --git a/intern/cycles/render/buffers.h b/intern/cycles/render/buffers.h
index 06b6094e6c9..425400a2c08 100644
--- a/intern/cycles/render/buffers.h
+++ b/intern/cycles/render/buffers.h
@@ -92,7 +92,7 @@ class RenderBuffers {
const string &name, float exposure, int sample, int components, float *pixels);
bool get_denoising_pass_rect(
int offset, float exposure, int sample, int components, float *pixels);
- bool set_pass_rect(PassType type, int components, float *pixels);
+ bool set_pass_rect(PassType type, int components, float *pixels, int samples);
};
/* Display Buffer
diff --git a/intern/cycles/render/film.cpp b/intern/cycles/render/film.cpp
index 7072fff4892..2da28222a7f 100644
--- a/intern/cycles/render/film.cpp
+++ b/intern/cycles/render/film.cpp
@@ -253,6 +253,8 @@ void Pass::add(PassType type, vector<Pass> &passes, const char *name)
case PASS_BAKE_PRIMITIVE:
case PASS_BAKE_DIFFERENTIAL:
pass.components = 4;
+ pass.exposure = false;
+ pass.filter = false;
break;
default:
assert(false);
generated by cgit v1.2.3 (git 2.25.1) at 2024-06-28 21:55:16 +0300