diff options
59 files changed, 1952 insertions, 354 deletions
diff --git a/CMakeLists.txt b/CMakeLists.txt index d2916283091..281aa52cae3 100644 --- a/CMakeLists.txt +++ b/CMakeLists.txt @@ -397,6 +397,7 @@ option(WITH_CYCLES "Enable Cycles Render Engine" ON) option(WITH_CYCLES_STANDALONE "Build Cycles standalone application" OFF) option(WITH_CYCLES_STANDALONE_GUI "Build Cycles standalone with GUI" OFF) option(WITH_CYCLES_OSL "Build Cycles with OSL support" ${_init_CYCLES_OSL}) +option(WITH_CYCLES_OPENSUBDIV "Build Cycles with OpenSubdiv support" ON) option(WITH_CYCLES_CUDA_BINARIES "Build Cycles CUDA binaries" OFF) set(CYCLES_CUDA_BINARIES_ARCH sm_20 sm_21 sm_30 sm_35 sm_37 sm_50 sm_52 CACHE STRING "CUDA architectures to build binaries for") mark_as_advanced(CYCLES_CUDA_BINARIES_ARCH) @@ -1101,7 +1102,6 @@ if(UNIX AND NOT APPLE) endif() if(WITH_ALEMBIC) - set(ALEMBIC_ROOT_DIR ${LIBDIR}/alembic) find_package_wrapper(Alembic) if(WITH_ALEMBIC_HDF5) @@ -1350,10 +1350,8 @@ elseif(WIN32) # MSVC11 needs _ALLOW_KEYWORD_MACROS to build add_definitions(-D_ALLOW_KEYWORD_MACROS) - if(CMAKE_CL_64) - # We want to support Vista level ABI for x64 - add_definitions(-D_WIN32_WINNT=0x600) - endif() + # We want to support Vista level ABI + add_definitions(-D_WIN32_WINNT=0x600) # Make cmake find the msvc redistributables set(CMAKE_INSTALL_SYSTEM_RUNTIME_LIBS_SKIP TRUE) @@ -1685,18 +1683,11 @@ elseif(WIN32) endif() if(WITH_ALEMBIC) - set(ALEMBIC_ROOT_DIR ${LIBDIR}/alembic) - find_package(Alembic) - - if(WITH_ALEMBIC_HDF5) - set(HDF5_ROOT_DIR ${LIBDIR}/hdf5) - find_package(HDF5) - endif() - - if(NOT ALEMBIC_FOUND OR (WITH_ALEMBIC_HDF5 AND NOT HDF5_FOUND)) - set(WITH_ALEMBIC OFF) - set(WITH_ALEMBIC_HDF5 OFF) - endif() + set(ALEMBIC ${LIBDIR}/alembic) + set(ALEMBIC_INCLUDE_DIR ${ALEMBIC}/include) + set(ALEMBIC_INCLUDE_DIRS ${ALEMBIC_INCLUDE_DIR}) + set(ALEMBIC_LIBPATH ${ALEMBIC}/lib) + set(ALEMBIC_LIBRARIES optimized alembic debug alembic_d) endif() if(WITH_MOD_CLOTH_ELTOPO) @@ -1999,14 +1990,16 @@ elseif(WIN32) endif() if(WITH_ALEMBIC) - set(ALEMBIC_ROOT_DIR ${LIBDIR}/alembic) + # TODO(sergey): For until someone drops by and compiles libraries for + # MinGW we allow users to compile their own Alembic library and use + # that via find_package(), + # + # Once precompiled libraries are there we'll use hardcoded locations. find_package_wrapper(Alembic) - if(WITH_ALEMBIC_HDF5) set(HDF5_ROOT_DIR ${LIBDIR}/hdf5) find_package_wrapper(HDF5) endif() - if(NOT ALEMBIC_FOUND OR (WITH_ALEMBIC_HDF5 AND NOT HDF5_FOUND)) set(WITH_ALEMBIC OFF) set(WITH_ALEMBIC_HDF5 OFF) @@ -2098,17 +2091,11 @@ elseif(APPLE) endif() if(WITH_ALEMBIC) - set(ALEMBIC_ROOT_DIR ${LIBDIR}/alembic) - find_package(Alembic) - if(WITH_ALEMBIC_HDF5) - set(HDF5_ROOT_DIR ${LIBDIR}/hdf5) - find_package(HDF5) - endif() - - if(NOT ALEMBIC_FOUND OR (WITH_ALEMBIC_HDF5 AND NOT HDF5_FOUND)) - set(WITH_ALEMBIC OFF) - set(WITH_ALEMBIC_HDF5 OFF) - endif() + set(ALEMBIC ${LIBDIR}/alembic) + set(ALEMBIC_INCLUDE_DIR ${ALEMBIC}/include) + set(ALEMBIC_INCLUDE_DIRS ${ALEMBIC_INCLUDE_DIR}) + set(ALEMBIC_LIBPATH ${ALEMBIC}/lib) + set(ALEMBIC_LIBRARIES Alembic) endif() if(WITH_OPENSUBDIV) @@ -2520,6 +2507,11 @@ if(WITH_CYCLES) ) endif() endif() + + if(WITH_CYCLES_OPENSUBDIV AND NOT WITH_OPENSUBDIV) + message(STATUS "WITH_CYCLES_OPENSUBDIV requires WITH_OPENSUBDIV to be ON, turning OFF") + set(WITH_CYCLES_OPENSUBDIV OFF) + endif() endif() if(WITH_INTERNATIONAL) @@ -2572,8 +2564,6 @@ else() endif() unset(_SYSTEM_BIG_ENDIAN) endif() - - if(WITH_IMAGE_OPENJPEG) if(WITH_SYSTEM_OPENJPEG) # dealt with above diff --git a/intern/cycles/CMakeLists.txt b/intern/cycles/CMakeLists.txt index 3b410b2a1e1..97854a88e84 100644 --- a/intern/cycles/CMakeLists.txt +++ b/intern/cycles/CMakeLists.txt @@ -146,6 +146,14 @@ if(WITH_CYCLES_OSL) ) endif() +if(WITH_CYCLES_OPENSUBDIV) + add_definitions(-DWITH_OPENSUBDIV) + include_directories( + SYSTEM + ${OPENSUBDIV_INCLUDE_DIR} + ) +endif() + set(WITH_CYCLES_DEVICE_OPENCL TRUE) set(WITH_CYCLES_DEVICE_CUDA TRUE) set(WITH_CYCLES_DEVICE_MULTI TRUE) diff --git a/intern/cycles/blender/addon/properties.py b/intern/cycles/blender/addon/properties.py index ab798373d9d..17defc5c0fa 100644 --- a/intern/cycles/blender/addon/properties.py +++ b/intern/cycles/blender/addon/properties.py @@ -775,6 +775,13 @@ class CyclesMaterialSettings(bpy.types.PropertyGroup): default='LINEAR', ) + cls.displacement_method = EnumProperty( + name="Displacement Method", + description="Method to use for the displacement", + items=enum_displacement_methods, + default='BUMP', + ) + @classmethod def unregister(cls): del bpy.types.Material.cycles @@ -952,13 +959,6 @@ class CyclesMeshSettings(bpy.types.PropertyGroup): type=cls, ) - cls.displacement_method = EnumProperty( - name="Displacement Method", - description="Method to use for the displacement", - items=enum_displacement_methods, - default='BUMP', - ) - @classmethod def unregister(cls): del bpy.types.Mesh.cycles diff --git a/intern/cycles/blender/addon/ui.py b/intern/cycles/blender/addon/ui.py index 65937908a67..3de309a34ea 100644 --- a/intern/cycles/blender/addon/ui.py +++ b/intern/cycles/blender/addon/ui.py @@ -674,40 +674,6 @@ class Cycles_PT_context_material(CyclesButtonsPanel, Panel): split.separator() -class Cycles_PT_mesh_displacement(CyclesButtonsPanel, Panel): - bl_label = "Displacement" - bl_context = "data" - - @classmethod - def poll(cls, context): - if CyclesButtonsPanel.poll(context): - if context.mesh or context.curve or context.meta_ball: - if context.scene.cycles.feature_set == 'EXPERIMENTAL': - return True - - return False - - def draw(self, context): - layout = self.layout - - mesh = context.mesh - curve = context.curve - mball = context.meta_ball - - if mesh: - cdata = mesh.cycles - elif curve: - cdata = curve.cycles - elif mball: - cdata = mball.cycles - - split = layout.split() - - col = split.column() - sub = col.column(align=True) - sub.label(text="Displacement:") - sub.prop(cdata, "displacement_method", text="") - class CyclesObject_PT_motion_blur(CyclesButtonsPanel, Panel): bl_label = "Motion Blur" bl_context = "object" @@ -1219,6 +1185,11 @@ class CyclesMaterial_PT_settings(CyclesButtonsPanel, Panel): col.prop(cmat, "sample_as_light", text="Multiple Importance") col.prop(cmat, "use_transparent_shadow") + if context.scene.cycles.feature_set == 'EXPERIMENTAL': + col.separator() + col.label(text="Displacement:") + col.prop(cmat, "displacement_method", text="") + col = split.column() col.label(text="Volume:") sub = col.column() diff --git a/intern/cycles/blender/blender_mesh.cpp b/intern/cycles/blender/blender_mesh.cpp index 74fd4cb44a0..6dc26c2981b 100644 --- a/intern/cycles/blender/blender_mesh.cpp +++ b/intern/cycles/blender/blender_mesh.cpp @@ -409,7 +409,8 @@ static void attr_create_uv_map(Scene *scene, BL::Mesh& b_mesh, const vector<int>& nverts, const vector<int>& face_flags, - bool subdivision) + bool subdivision, + bool subdivide_uvs) { if(subdivision) { BL::Mesh::uv_layers_iterator l; @@ -429,6 +430,10 @@ static void attr_create_uv_map(Scene *scene, else attr = mesh->subd_attributes.add(name, TypeDesc::TypePoint, ATTR_ELEMENT_CORNER); + if(subdivide_uvs) { + attr->flags |= ATTR_SUBDIVIDED; + } + BL::Mesh::polygons_iterator p; float3 *fdata = attr->data_float3(); @@ -592,7 +597,8 @@ static void create_mesh(Scene *scene, Mesh *mesh, BL::Mesh& b_mesh, const vector<Shader*>& used_shaders, - bool subdivision=false) + bool subdivision=false, + bool subdivide_uvs=true) { /* count vertices and faces */ int numverts = b_mesh.vertices.length(); @@ -638,6 +644,7 @@ static void create_mesh(Scene *scene, /* create generated coordinates from undeformed coordinates */ if(mesh->need_attribute(scene, ATTR_STD_GENERATED)) { Attribute *attr = attributes.add(ATTR_STD_GENERATED); + attr->flags |= ATTR_SUBDIVIDED; float3 loc, size; mesh_texture_space(b_mesh, loc, size); @@ -746,7 +753,7 @@ static void create_mesh(Scene *scene, * The calculate functions will check whether they're needed or not. */ attr_create_vertex_color(scene, mesh, b_mesh, nverts, face_flags, subdivision); - attr_create_uv_map(scene, mesh, b_mesh, nverts, face_flags, subdivision); + attr_create_uv_map(scene, mesh, b_mesh, nverts, face_flags, subdivision, subdivide_uvs); /* for volume objects, create a matrix to transform from object space to * mesh texture space. this does not work with deformations but that can @@ -770,8 +777,35 @@ static void create_subd_mesh(Scene *scene, float dicing_rate, int max_subdivisions) { - create_mesh(scene, mesh, b_mesh, used_shaders, true); + BL::SubsurfModifier subsurf_mod(b_ob.modifiers[b_ob.modifiers.length()-1]); + bool subdivide_uvs = subsurf_mod.use_subsurf_uv(); + + create_mesh(scene, mesh, b_mesh, used_shaders, true, subdivide_uvs); + + /* export creases */ + size_t num_creases = 0; + BL::Mesh::edges_iterator e; + + for(b_mesh.edges.begin(e); e != b_mesh.edges.end(); ++e) { + if(e->crease() != 0.0f) { + num_creases++; + } + } + + mesh->subd_creases.resize(num_creases); + Mesh::SubdEdgeCrease* crease = mesh->subd_creases.data(); + for(b_mesh.edges.begin(e); e != b_mesh.edges.end(); ++e) { + if(e->crease() != 0.0f) { + crease->v[0] = e->vertices()[0]; + crease->v[1] = e->vertices()[1]; + crease->crease = e->crease(); + + crease++; + } + } + + /* set subd params */ SubdParams sdparams(mesh); PointerRNA cobj = RNA_pointer_get(&b_ob.ptr, "cycles"); @@ -903,8 +937,6 @@ Mesh *BlenderSync::sync_mesh(BL::Object& b_ob, mesh_synced.insert(mesh); /* create derived mesh */ - PointerRNA cmesh = RNA_pointer_get(&b_ob_data.ptr, "cycles"); - array<int> oldtriangle = mesh->triangles; /* compares curve_keys rather than strands in order to handle quick hair @@ -936,7 +968,7 @@ Mesh *BlenderSync::sync_mesh(BL::Object& b_ob, BL::Modifier mod = b_ob.modifiers[b_ob.modifiers.length()-1]; bool enabled = preview ? mod.show_viewport() : mod.show_render(); - if(enabled && mod.type() == BL::Modifier::type_SUBSURF && RNA_int_get(&cobj, "use_adaptive_subdivision")) { + if(enabled && mod.type() == BL::Modifier::type_SUBSURF && RNA_boolean_get(&cobj, "use_adaptive_subdivision")) { BL::SubsurfModifier subsurf(mod); if(subsurf.subdivision_type() == BL::SubsurfModifier::subdivision_type_CATMULL_CLARK) { @@ -974,21 +1006,6 @@ Mesh *BlenderSync::sync_mesh(BL::Object& b_ob, } mesh->geometry_flags = requested_geometry_flags; - /* displacement method */ - if(cmesh.data) { - const int method = get_enum(cmesh, - "displacement_method", - Mesh::DISPLACE_NUM_METHODS, - Mesh::DISPLACE_BUMP); - - if(method == 0 || !experimental) - mesh->displacement_method = Mesh::DISPLACE_BUMP; - else if(method == 1) - mesh->displacement_method = Mesh::DISPLACE_TRUE; - else - mesh->displacement_method = Mesh::DISPLACE_BOTH; - } - /* fluid motion */ sync_mesh_fluid_motion(b_ob, scene, mesh); diff --git a/intern/cycles/blender/blender_object.cpp b/intern/cycles/blender/blender_object.cpp index 9d91d9b0a63..8e798d18647 100644 --- a/intern/cycles/blender/blender_object.cpp +++ b/intern/cycles/blender/blender_object.cpp @@ -329,16 +329,18 @@ Object *BlenderSync::sync_object(BL::Object& b_parent, /* object transformation */ if(tfm != object->tfm) { VLOG(1) << "Object " << b_ob.name() << " motion detected."; - if(motion_time == -1.0f) { - object->motion.pre = tfm; - object->use_motion = true; - } - else if(motion_time == 1.0f) { - object->motion.post = tfm; + if(motion_time == -1.0f || motion_time == 1.0f) { object->use_motion = true; } } + if(motion_time == -1.0f) { + object->motion.pre = tfm; + } + else if(motion_time == 1.0f) { + object->motion.post = tfm; + } + /* mesh deformation */ if(object->mesh) sync_mesh_motion(b_ob, object, motion_time); diff --git a/intern/cycles/blender/blender_shader.cpp b/intern/cycles/blender/blender_shader.cpp index 2fe8ee90334..534bc6cc897 100644 --- a/intern/cycles/blender/blender_shader.cpp +++ b/intern/cycles/blender/blender_shader.cpp @@ -64,6 +64,14 @@ static VolumeInterpolation get_volume_interpolation(PointerRNA& ptr) VOLUME_INTERPOLATION_LINEAR); } +static DisplacementMethod get_displacement_method(PointerRNA& ptr) +{ + return (DisplacementMethod)get_enum(ptr, + "displacement_method", + DISPLACE_NUM_METHODS, + DISPLACE_BUMP); +} + static int validate_enum_value(int value, int num_values, int default_value) { if(value >= num_values) { @@ -1182,6 +1190,7 @@ void BlenderSync::sync_materials(bool update_all) shader->heterogeneous_volume = !get_boolean(cmat, "homogeneous_volume"); shader->volume_sampling_method = get_volume_sampling(cmat); shader->volume_interpolation_method = get_volume_interpolation(cmat); + shader->displacement_method = (experimental) ? get_displacement_method(cmat) : DISPLACE_BUMP; shader->set_graph(graph); shader->tag_update(scene); diff --git a/intern/cycles/device/device_cuda.cpp b/intern/cycles/device/device_cuda.cpp index 6a511ea7316..a85f34082db 100644 --- a/intern/cycles/device/device_cuda.cpp +++ b/intern/cycles/device/device_cuda.cpp @@ -297,7 +297,7 @@ public: cuda_error_message("CUDA nvcc compiler version could not be parsed."); return false; } - if(cuda_version < 60) { + if(cuda_version < 75) { printf("Unsupported CUDA version %d.%d detected, " "you need CUDA 7.5 or newer.\n", major, minor); diff --git a/intern/cycles/kernel/CMakeLists.txt b/intern/cycles/kernel/CMakeLists.txt index 42298a0811d..7bef247d3bd 100644 --- a/intern/cycles/kernel/CMakeLists.txt +++ b/intern/cycles/kernel/CMakeLists.txt @@ -162,6 +162,7 @@ set(SRC_GEOM_HEADERS geom/geom_motion_curve.h geom/geom_motion_triangle.h geom/geom_object.h + geom/geom_patch.h geom/geom_primitive.h geom/geom_subd_triangle.h geom/geom_triangle.h diff --git a/intern/cycles/kernel/geom/geom.h b/intern/cycles/kernel/geom/geom.h index 493afdc4f62..11548324e18 100644 --- a/intern/cycles/kernel/geom/geom.h +++ b/intern/cycles/kernel/geom/geom.h @@ -17,6 +17,7 @@ #include "geom_attribute.h" #include "geom_object.h" +#include "geom_patch.h" #include "geom_triangle.h" #include "geom_subd_triangle.h" #include "geom_triangle_intersect.h" diff --git a/intern/cycles/kernel/geom/geom_attribute.h b/intern/cycles/kernel/geom/geom_attribute.h index e036c2752d7..8604d30ad34 100644 --- a/intern/cycles/kernel/geom/geom_attribute.h +++ b/intern/cycles/kernel/geom/geom_attribute.h @@ -45,7 +45,7 @@ ccl_device_inline uint attribute_primitive_type(KernelGlobals *kg, const ShaderD ccl_device_inline AttributeDescriptor attribute_not_found() { - const AttributeDescriptor desc = {ATTR_ELEMENT_NONE, (NodeAttributeType)0, ATTR_STD_NOT_FOUND}; + const AttributeDescriptor desc = {ATTR_ELEMENT_NONE, (NodeAttributeType)0, 0, ATTR_STD_NOT_FOUND}; return desc; } @@ -79,7 +79,8 @@ ccl_device_inline AttributeDescriptor find_attribute(KernelGlobals *kg, const Sh /* return result */ desc.offset = (attr_map.y == ATTR_ELEMENT_NONE) ? (int)ATTR_STD_NOT_FOUND : (int)attr_map.z; - desc.type = (NodeAttributeType)attr_map.w; + desc.type = (NodeAttributeType)(attr_map.w & 0xff); + desc.flags = (AttributeFlag)(attr_map.w >> 8); return desc; } diff --git a/intern/cycles/kernel/geom/geom_object.h b/intern/cycles/kernel/geom/geom_object.h index c0d15a95954..883c5dc100d 100644 --- a/intern/cycles/kernel/geom/geom_object.h +++ b/intern/cycles/kernel/geom/geom_object.h @@ -292,6 +292,18 @@ ccl_device_inline void object_motion_info(KernelGlobals *kg, int object, int *nu *numverts = __float_as_int(f.w); } +/* Offset to an objects patch map */ + +ccl_device_inline uint object_patch_map_offset(KernelGlobals *kg, int object) +{ + if(object == OBJECT_NONE) + return 0; + + int offset = object*OBJECT_SIZE + 11; + float4 f = kernel_tex_fetch(__objects, offset); + return __float_as_uint(f.x); +} + /* Pass ID for shader */ ccl_device int shader_pass_id(KernelGlobals *kg, const ShaderData *sd) diff --git a/intern/cycles/kernel/geom/geom_patch.h b/intern/cycles/kernel/geom/geom_patch.h new file mode 100644 index 00000000000..6a0ff5a4a04 --- /dev/null +++ b/intern/cycles/kernel/geom/geom_patch.h @@ -0,0 +1,343 @@ +/* + * Based on code from OpenSubdiv released under this license: + * + * Copyright 2013 Pixar + * + * Licensed under the Apache License, Version 2.0 (the "Apache License") + * with the following modification; you may not use this file except in + * compliance with the Apache License and the following modification to it: + * Section 6. Trademarks. is deleted and replaced with: + * + * 6. Trademarks. This License does not grant permission to use the trade + * names, trademarks, service marks, or product names of the Licensor + * and its affiliates, except as required to comply with Section 4(c) of + * the License and to reproduce the content of the NOTICE file. + * + * You may obtain a copy of the Apache License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the Apache License with the above modification is + * distributed on an "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY + * KIND, either express or implied. See the Apache License for the specific + * language governing permissions and limitations under the Apache License. + * + */ + +CCL_NAMESPACE_BEGIN + +typedef struct PatchHandle { + int array_index, patch_index, vert_index; +} PatchHandle; + +ccl_device_inline int patch_map_resolve_quadrant(float median, float *u, float *v) +{ + int quadrant = -1; + + if(*u < median) { + if(*v < median) { + quadrant = 0; + } + else { + quadrant = 1; + *v -= median; + } + } + else { + if(*v < median) { + quadrant = 3; + } + else { + quadrant = 2; + *v -= median; + } + *u -= median; + } + + return quadrant; +} + +/* retrieve PatchHandle from patch coords */ + +ccl_device_inline PatchHandle patch_map_find_patch(KernelGlobals *kg, int object, int patch, float u, float v) +{ + PatchHandle handle; + + kernel_assert((u >= 0.0f) && (u <= 1.0f) && (v >= 0.0f) && (v <= 1.0f)); + + int node = (object_patch_map_offset(kg, object) + patch)/2; + float median = 0.5f; + + for(int depth = 0; depth < 0xff; depth++) { + float delta = median * 0.5f; + + int quadrant = patch_map_resolve_quadrant(median, &u, &v); + kernel_assert(quadrant >= 0); + + uint child = kernel_tex_fetch(__patches, node + quadrant); + + /* is the quadrant a hole? */ + if(!(child & PATCH_MAP_NODE_IS_SET)) { + handle.array_index = -1; + return handle; + } + + uint index = child & PATCH_MAP_NODE_INDEX_MASK; + + if(child & PATCH_MAP_NODE_IS_LEAF) { + handle.array_index = kernel_tex_fetch(__patches, index + 0); + handle.patch_index = kernel_tex_fetch(__patches, index + 1); + handle.vert_index = kernel_tex_fetch(__patches, index + 2); + + return handle; + } else { + node = index; + } + + median = delta; + } + + /* no leaf found */ + kernel_assert(0); + + handle.array_index = -1; + return handle; +} + +ccl_device_inline void patch_eval_bspline_weights(float t, float *point, float *deriv) +{ + /* The four uniform cubic B-Spline basis functions evaluated at t */ + float inv_6 = 1.0f / 6.0f; + + float t2 = t * t; + float t3 = t * t2; + + point[0] = inv_6 * (1.0f - 3.0f*(t - t2) - t3); + point[1] = inv_6 * (4.0f - 6.0f*t2 + 3.0f*t3); + point[2] = inv_6 * (1.0f + 3.0f*(t + t2 - t3)); + point[3] = inv_6 * t3; + + /* Derivatives of the above four basis functions at t */ + deriv[0] = -0.5f*t2 + t - 0.5f; + deriv[1] = 1.5f*t2 - 2.0f*t; + deriv[2] = -1.5f*t2 + t + 0.5f; + deriv[3] = 0.5f*t2; +} + +ccl_device_inline void patch_eval_adjust_boundary_weights(uint bits, float *s, float *t) +{ + int boundary = ((bits >> 8) & 0xf); + + if(boundary & 1) { + t[2] -= t[0]; + t[1] += 2*t[0]; + t[0] = 0; + } + + if(boundary & 2) { + s[1] -= s[3]; + s[2] += 2*s[3]; + s[3] = 0; + } + + if(boundary & 4) { + t[1] -= t[3]; + t[2] += 2*t[3]; + t[3] = 0; + } + + if(boundary & 8) { + s[2] -= s[0]; + s[1] += 2*s[0]; + s[0] = 0; + } +} + +ccl_device_inline int patch_eval_depth(uint patch_bits) +{ + return (patch_bits & 0xf); +} + +ccl_device_inline float patch_eval_param_fraction(uint patch_bits) +{ + bool non_quad_root = (patch_bits >> 4) & 0x1; + int depth = patch_eval_depth(patch_bits); + + if(non_quad_root) { + return 1.0f / (float)(1 << (depth-1)); + } + else { + return 1.0f / (float)(1 << depth); + } +} + +ccl_device_inline void patch_eval_normalize_coords(uint patch_bits, float *u, float *v) +{ + float frac = patch_eval_param_fraction(patch_bits); + + int iu = (patch_bits >> 22) & 0x3ff; + int iv = (patch_bits >> 12) & 0x3ff; + + /* top left corner */ + float pu = (float)iu*frac; + float pv = (float)iv*frac; + + /* normalize uv coordinates */ + *u = (*u - pu) / frac; + *v = (*v - pv) / frac; +} + +/* retrieve patch control indices */ + +ccl_device_inline int patch_eval_indices(KernelGlobals *kg, const PatchHandle *handle, int channel, + int indices[PATCH_MAX_CONTROL_VERTS]) +{ + int index_base = kernel_tex_fetch(__patches, handle->array_index + 2) + handle->vert_index; + + /* XXX: regular patches only */ + for(int i = 0; i < 16; i++) { + indices[i] = kernel_tex_fetch(__patches, index_base + i); + } + + return 16; +} + +/* evaluate patch basis functions */ + +ccl_device_inline void patch_eval_basis(KernelGlobals *kg, const PatchHandle *handle, float u, float v, + float weights[PATCH_MAX_CONTROL_VERTS], + float weights_du[PATCH_MAX_CONTROL_VERTS], + float weights_dv[PATCH_MAX_CONTROL_VERTS]) +{ + uint patch_bits = kernel_tex_fetch(__patches, handle->patch_index + 1); /* read patch param */ + float d_scale = 1 << patch_eval_depth(patch_bits); + + bool non_quad_root = (patch_bits >> 4) & 0x1; + if(non_quad_root) { + d_scale *= 0.5f; + } + + patch_eval_normalize_coords(patch_bits, &u, &v); + + /* XXX: regular patches only for now. */ + + float s[4], t[4], ds[4], dt[4]; + + patch_eval_bspline_weights(u, s, ds); + patch_eval_bspline_weights(v, t, dt); + + patch_eval_adjust_boundary_weights(patch_bits, s, t); + patch_eval_adjust_boundary_weights(patch_bits, ds, dt); + + for(int k = 0; k < 4; k++) { + for(int l = 0; l < 4; l++) { + weights[4*k+l] = s[l] * t[k]; + weights_du[4*k+l] = ds[l] * t[k] * d_scale; + weights_dv[4*k+l] = s[l] * dt[k] * d_scale; + } + } +} + +/* generic function for evaluating indices and weights from patch coords */ + +ccl_device_inline int patch_eval_control_verts(KernelGlobals *kg, int object, int patch, float u, float v, int channel, + int indices[PATCH_MAX_CONTROL_VERTS], + float weights[PATCH_MAX_CONTROL_VERTS], + float weights_du[PATCH_MAX_CONTROL_VERTS], + float weights_dv[PATCH_MAX_CONTROL_VERTS]) +{ + PatchHandle handle = patch_map_find_patch(kg, object, patch, u, v); + kernel_assert(handle.array_index >= 0); + + int num_control = patch_eval_indices(kg, &handle, channel, indices); + patch_eval_basis(kg, &handle, u, v, weights, weights_du, weights_dv); + + return num_control; +} + +/* functions for evaluating attributes on patches */ + +ccl_device float patch_eval_float(KernelGlobals *kg, const ShaderData *sd, int offset, + int patch, float u, float v, int channel, + float *du, float* dv) +{ + int indices[PATCH_MAX_CONTROL_VERTS]; + float weights[PATCH_MAX_CONTROL_VERTS]; + float weights_du[PATCH_MAX_CONTROL_VERTS]; + float weights_dv[PATCH_MAX_CONTROL_VERTS]; + + int num_control = patch_eval_control_verts(kg, ccl_fetch(sd, object), patch, u, v, channel, + indices, weights, weights_du, weights_dv); + + float val = 0.0f; + if(du) *du = 0.0f; + if(dv) *dv = 0.0f; + + for(int i = 0; i < num_control; i++) { + float v = kernel_tex_fetch(__attributes_float, offset + indices[i]); + + val += v * weights[i]; + if(du) *du += v * weights_du[i]; + if(dv) *dv += v * weights_dv[i]; + } + + return val; +} + +ccl_device float3 patch_eval_float3(KernelGlobals *kg, const ShaderData *sd, int offset, + int patch, float u, float v, int channel, + float3 *du, float3 *dv) +{ + int indices[PATCH_MAX_CONTROL_VERTS]; + float weights[PATCH_MAX_CONTROL_VERTS]; + float weights_du[PATCH_MAX_CONTROL_VERTS]; + float weights_dv[PATCH_MAX_CONTROL_VERTS]; + + int num_control = patch_eval_control_verts(kg, ccl_fetch(sd, object), patch, u, v, channel, + indices, weights, weights_du, weights_dv); + + float3 val = make_float3(0.0f, 0.0f, 0.0f); + if(du) *du = make_float3(0.0f, 0.0f, 0.0f); + if(dv) *dv = make_float3(0.0f, 0.0f, 0.0f); + + for(int i = 0; i < num_control; i++) { + float3 v = float4_to_float3(kernel_tex_fetch(__attributes_float3, offset + indices[i])); + + val += v * weights[i]; + if(du) *du += v * weights_du[i]; + if(dv) *dv += v * weights_dv[i]; + } + + return val; +} + +ccl_device float3 patch_eval_uchar4(KernelGlobals *kg, const ShaderData *sd, int offset, + int patch, float u, float v, int channel, + float3 *du, float3 *dv) +{ + int indices[PATCH_MAX_CONTROL_VERTS]; + float weights[PATCH_MAX_CONTROL_VERTS]; + float weights_du[PATCH_MAX_CONTROL_VERTS]; + float weights_dv[PATCH_MAX_CONTROL_VERTS]; + + int num_control = patch_eval_control_verts(kg, ccl_fetch(sd, object), patch, u, v, channel, + indices, weights, weights_du, weights_dv); + + float3 val = make_float3(0.0f, 0.0f, 0.0f); + if(du) *du = make_float3(0.0f, 0.0f, 0.0f); + if(dv) *dv = make_float3(0.0f, 0.0f, 0.0f); + + for(int i = 0; i < num_control; i++) { + float3 v = color_byte_to_float(kernel_tex_fetch(__attributes_uchar4, offset + indices[i])); + + val += v * weights[i]; + if(du) *du += v * weights_du[i]; + if(dv) *dv += v * weights_dv[i]; + } + + return val; +} + +CCL_NAMESPACE_END + diff --git a/intern/cycles/kernel/geom/geom_subd_triangle.h b/intern/cycles/kernel/geom/geom_subd_triangle.h index 79a2ce5cc70..fccacf435f9 100644 --- a/intern/cycles/kernel/geom/geom_subd_triangle.h +++ b/intern/cycles/kernel/geom/geom_subd_triangle.h @@ -97,11 +97,54 @@ ccl_device_inline void subd_triangle_patch_corners(KernelGlobals *kg, int patch, /* Reading attributes on various subdivision triangle elements */ -ccl_device float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderData *sd, const AttributeDescriptor desc, float *dx, float *dy) +ccl_device_noinline float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderData *sd, const AttributeDescriptor desc, float *dx, float *dy) { int patch = subd_triangle_patch(kg, sd); - if(desc.element == ATTR_ELEMENT_FACE) { + if(desc.flags & ATTR_SUBDIVIDED) { + float2 uv[3]; + subd_triangle_patch_uv(kg, sd, uv); + + float2 dpdu = uv[0] - uv[2]; + float2 dpdv = uv[1] - uv[2]; + + /* p is [s, t] */ + float2 p = dpdu * ccl_fetch(sd, u) + dpdv * ccl_fetch(sd, v) + uv[2]; + + float a, dads, dadt; + a = patch_eval_float(kg, sd, desc.offset, patch, p.x, p.y, 0, &dads, &dadt); + +#ifdef __RAY_DIFFERENTIALS__ + if(dx || dy) { + float dsdu = dpdu.x; + float dtdu = dpdu.y; + float dsdv = dpdv.x; + float dtdv = dpdv.y; + + if(dx) { + float dudx = ccl_fetch(sd, du).dx; + float dvdx = ccl_fetch(sd, dv).dx; + + float dsdx = dsdu*dudx + dsdv*dvdx; + float dtdx = dtdu*dudx + dtdv*dvdx; + + *dx = dads*dsdx + dadt*dtdx; + } + if(dy) { + float dudy = ccl_fetch(sd, du).dy; + float dvdy = ccl_fetch(sd, dv).dy; + + float dsdy = dsdu*dudy + dsdv*dvdy; + float dtdy = dtdu*dudy + dtdv*dvdy; + + *dy = dads*dsdy + dadt*dtdy; + } + } +#endif + + return a; + } + else if(desc.element == ATTR_ELEMENT_FACE) { if(dx) *dx = 0.0f; if(dy) *dy = 0.0f; @@ -110,9 +153,8 @@ ccl_device float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderDa else if(desc.element == ATTR_ELEMENT_VERTEX || desc.element == ATTR_ELEMENT_VERTEX_MOTION) { float2 uv[3]; subd_triangle_patch_uv(kg, sd, uv); - uint4 v = subd_triangle_patch_indices(kg, patch); - float a, b, c; + uint4 v = subd_triangle_patch_indices(kg, patch); float f0 = kernel_tex_fetch(__attributes_float, desc.offset + v.x); float f1 = kernel_tex_fetch(__attributes_float, desc.offset + v.y); @@ -124,9 +166,9 @@ ccl_device float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderDa f3 = (f3+f0)*0.5f; } - a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); - b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); - c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); + float a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); + float b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); + float c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); #ifdef __RAY_DIFFERENTIALS__ if(dx) *dx = ccl_fetch(sd, du).dx*a + ccl_fetch(sd, dv).dx*b - (ccl_fetch(sd, du).dx + ccl_fetch(sd, dv).dx)*c; @@ -136,13 +178,11 @@ ccl_device float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderDa return ccl_fetch(sd, u)*a + ccl_fetch(sd, v)*b + (1.0f - ccl_fetch(sd, u) - ccl_fetch(sd, v))*c; } else if(desc.element == ATTR_ELEMENT_CORNER) { - int corners[4]; - subd_triangle_patch_corners(kg, patch, corners); - float2 uv[3]; subd_triangle_patch_uv(kg, sd, uv); - float a, b, c; + int corners[4]; + subd_triangle_patch_corners(kg, patch, corners); float f0 = kernel_tex_fetch(__attributes_float, corners[0] + desc.offset); float f1 = kernel_tex_fetch(__attributes_float, corners[1] + desc.offset); @@ -154,9 +194,9 @@ ccl_device float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderDa f3 = (f3+f0)*0.5f; } - a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); - b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); - c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); + float a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); + float b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); + float c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); #ifdef __RAY_DIFFERENTIALS__ if(dx) *dx = ccl_fetch(sd, du).dx*a + ccl_fetch(sd, dv).dx*b - (ccl_fetch(sd, du).dx + ccl_fetch(sd, dv).dx)*c; @@ -173,11 +213,60 @@ ccl_device float subd_triangle_attribute_float(KernelGlobals *kg, const ShaderDa } } -ccl_device float3 subd_triangle_attribute_float3(KernelGlobals *kg, const ShaderData *sd, const AttributeDescriptor desc, float3 *dx, float3 *dy) +ccl_device_noinline float3 subd_triangle_attribute_float3(KernelGlobals *kg, const ShaderData *sd, const AttributeDescriptor desc, float3 *dx, float3 *dy) { int patch = subd_triangle_patch(kg, sd); - if(desc.element == ATTR_ELEMENT_FACE) { + if(desc.flags & ATTR_SUBDIVIDED) { + float2 uv[3]; + subd_triangle_patch_uv(kg, sd, uv); + + float2 dpdu = uv[0] - uv[2]; + float2 dpdv = uv[1] - uv[2]; + + /* p is [s, t] */ + float2 p = dpdu * ccl_fetch(sd, u) + dpdv * ccl_fetch(sd, v) + uv[2]; + + float3 a, dads, dadt; + + if(desc.element == ATTR_ELEMENT_CORNER_BYTE) { + a = patch_eval_uchar4(kg, sd, desc.offset, patch, p.x, p.y, 0, &dads, &dadt); + } + else { + a = patch_eval_float3(kg, sd, desc.offset, patch, p.x, p.y, 0, &dads, &dadt); + } + +#ifdef __RAY_DIFFERENTIALS__ + if(dx || dy) { + float dsdu = dpdu.x; + float dtdu = dpdu.y; + float dsdv = dpdv.x; + float dtdv = dpdv.y; + + if(dx) { + float dudx = ccl_fetch(sd, du).dx; + float dvdx = ccl_fetch(sd, dv).dx; + + float dsdx = dsdu*dudx + dsdv*dvdx; + float dtdx = dtdu*dudx + dtdv*dvdx; + + *dx = dads*dsdx + dadt*dtdx; + } + if(dy) { + float dudy = ccl_fetch(sd, du).dy; + float dvdy = ccl_fetch(sd, dv).dy; + + float dsdy = dsdu*dudy + dsdv*dvdy; + float dtdy = dtdu*dudy + dtdv*dvdy; + + *dy = dads*dsdy + dadt*dtdy; + } + } +#endif + + return a; + } + else if(desc.element == ATTR_ELEMENT_FACE) { if(dx) *dx = make_float3(0.0f, 0.0f, 0.0f); if(dy) *dy = make_float3(0.0f, 0.0f, 0.0f); @@ -186,9 +275,8 @@ ccl_device float3 subd_triangle_attribute_float3(KernelGlobals *kg, const Shader else if(desc.element == ATTR_ELEMENT_VERTEX || desc.element == ATTR_ELEMENT_VERTEX_MOTION) { float2 uv[3]; subd_triangle_patch_uv(kg, sd, uv); - uint4 v = subd_triangle_patch_indices(kg, patch); - float3 a, b, c; + uint4 v = subd_triangle_patch_indices(kg, patch); float3 f0 = float4_to_float3(kernel_tex_fetch(__attributes_float3, desc.offset + v.x)); float3 f1 = float4_to_float3(kernel_tex_fetch(__attributes_float3, desc.offset + v.y)); @@ -200,9 +288,9 @@ ccl_device float3 subd_triangle_attribute_float3(KernelGlobals *kg, const Shader f3 = (f3+f0)*0.5f; } - a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); - b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); - c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); + float3 a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); + float3 b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); + float3 c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); #ifdef __RAY_DIFFERENTIALS__ if(dx) *dx = ccl_fetch(sd, du).dx*a + ccl_fetch(sd, dv).dx*b - (ccl_fetch(sd, du).dx + ccl_fetch(sd, dv).dx)*c; @@ -212,13 +300,12 @@ ccl_device float3 subd_triangle_attribute_float3(KernelGlobals *kg, const Shader return ccl_fetch(sd, u)*a + ccl_fetch(sd, v)*b + (1.0f - ccl_fetch(sd, u) - ccl_fetch(sd, v))*c; } else if(desc.element == ATTR_ELEMENT_CORNER || desc.element == ATTR_ELEMENT_CORNER_BYTE) { - int corners[4]; - subd_triangle_patch_corners(kg, patch, corners); - float2 uv[3]; subd_triangle_patch_uv(kg, sd, uv); - float3 a, b, c; + int corners[4]; + subd_triangle_patch_corners(kg, patch, corners); + float3 f0, f1, f2, f3; if(desc.element == ATTR_ELEMENT_CORNER) { @@ -239,9 +326,9 @@ ccl_device float3 subd_triangle_attribute_float3(KernelGlobals *kg, const Shader f3 = (f3+f0)*0.5f; } - a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); - b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); - c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); + float3 a = mix(mix(f0, f1, uv[0].x), mix(f3, f2, uv[0].x), uv[0].y); + float3 b = mix(mix(f0, f1, uv[1].x), mix(f3, f2, uv[1].x), uv[1].y); + float3 c = mix(mix(f0, f1, uv[2].x), mix(f3, f2, uv[2].x), uv[2].y); #ifdef __RAY_DIFFERENTIALS__ if(dx) *dx = ccl_fetch(sd, du).dx*a + ccl_fetch(sd, dv).dx*b - (ccl_fetch(sd, du).dx + ccl_fetch(sd, dv).dx)*c; diff --git a/intern/cycles/kernel/kernel_compat_cuda.h b/intern/cycles/kernel/kernel_compat_cuda.h index a039b414006..063220b542e 100644 --- a/intern/cycles/kernel/kernel_compat_cuda.h +++ b/intern/cycles/kernel/kernel_compat_cuda.h @@ -47,6 +47,7 @@ #define ccl_may_alias #define ccl_addr_space #define ccl_restrict __restrict__ +#define ccl_align(n) __align__(n) /* No assert supported for CUDA */ diff --git a/intern/cycles/kernel/kernel_compat_opencl.h b/intern/cycles/kernel/kernel_compat_opencl.h index 8505cb85576..2ae89dde7c4 100644 --- a/intern/cycles/kernel/kernel_compat_opencl.h +++ b/intern/cycles/kernel/kernel_compat_opencl.h @@ -40,6 +40,7 @@ #define ccl_local __local #define ccl_private __private #define ccl_restrict restrict +#define ccl_align(n) __attribute__((aligned(n))) #ifdef __SPLIT_KERNEL__ # define ccl_addr_space __global diff --git a/intern/cycles/kernel/kernel_types.h b/intern/cycles/kernel/kernel_types.h index 57b116e1dd7..f3b10c21b9d 100644 --- a/intern/cycles/kernel/kernel_types.h +++ b/intern/cycles/kernel/kernel_types.h @@ -34,7 +34,7 @@ CCL_NAMESPACE_BEGIN /* constants */ -#define OBJECT_SIZE 11 +#define OBJECT_SIZE 12 #define OBJECT_VECTOR_SIZE 6 #define LIGHT_SIZE 5 #define FILTER_TABLE_SIZE 1024 @@ -624,9 +624,15 @@ typedef enum AttributeStandard { ATTR_STD_NOT_FOUND = ~0 } AttributeStandard; +typedef enum AttributeFlag { + ATTR_FINAL_SIZE = (1 << 0), + ATTR_SUBDIVIDED = (1 << 1), +} AttributeFlag; + typedef struct AttributeDescriptor { AttributeElement element; NodeAttributeType type; + uint flags; /* see enum AttributeFlag */ int offset; } AttributeDescriptor; @@ -650,23 +656,18 @@ typedef struct AttributeDescriptor { * ShaderClosure has a fixed size, and any extra space must be allocated * with closure_alloc_extra(). * - * float3 is 12 bytes on CUDA and 16 bytes on CPU/OpenCL, we set the data - * size to ensure ShaderClosure is 80 bytes total everywhere. */ + * We pad the struct to 80 bytes and ensure it is aligned to 16 bytes, which + * we assume to be the maximum required alignment for any struct. */ #define SHADER_CLOSURE_BASE \ float3 weight; \ ClosureType type; \ float sample_weight \ -typedef ccl_addr_space struct ShaderClosure { +typedef ccl_addr_space struct ccl_align(16) ShaderClosure { SHADER_CLOSURE_BASE; - /* pad to 80 bytes, data types are aligned to own size */ -#ifdef __KERNEL_CUDA__ - float data[15]; -#else - float data[14]; -#endif + float data[14]; /* pad to 80 bytes */ } ShaderClosure; /* Shader Context @@ -741,7 +742,7 @@ enum ShaderDataFlag { # define SD_THREAD (get_global_id(1) * get_global_size(0) + get_global_id(0)) # if defined(__SPLIT_KERNEL_AOS__) /* ShaderData is stored as an Array-of-Structures */ -# define ccl_soa_member(type, name) type soa_##name; +# define ccl_soa_member(type, name) type soa_##name # define ccl_fetch(s, t) (s[SD_THREAD].soa_##t) # define ccl_fetch_array(s, t, index) (&s[SD_THREAD].soa_##t[index]) # else @@ -749,7 +750,7 @@ enum ShaderDataFlag { # define SD_GLOBAL_SIZE (get_global_size(0) * get_global_size(1)) # define SD_FIELD_SIZE(t) sizeof(((struct ShaderData*)0)->t) # define SD_OFFSETOF(t) ((char*)(&((struct ShaderData*)0)->t) - (char*)0) -# define ccl_soa_member(type, name) type soa_##name; +# define ccl_soa_member(type, name) type soa_##name # define ccl_fetch(s, t) (((ShaderData*)((ccl_addr_space char*)s + SD_GLOBAL_SIZE * SD_OFFSETOF(soa_##t) + SD_FIELD_SIZE(soa_##t) * SD_THREAD - SD_OFFSETOF(soa_##t)))->soa_##t) # define ccl_fetch_array(s, t, index) (&ccl_fetch(s, t)[index]) # endif @@ -1245,6 +1246,16 @@ enum RayState { #define REMOVE_RAY_FLAG(ray_state, ray_index, flag) (ray_state[ray_index] = (ray_state[ray_index] & (~flag))) #define IS_FLAG(ray_state, ray_index, flag) (ray_state[ray_index] & flag) +/* Patches */ + +#define PATCH_MAX_CONTROL_VERTS 16 + +/* Patch map node flags */ + +#define PATCH_MAP_NODE_IS_SET (1 << 30) +#define PATCH_MAP_NODE_IS_LEAF (1 << 31) +#define PATCH_MAP_NODE_INDEX_MASK (~(PATCH_MAP_NODE_IS_SET | PATCH_MAP_NODE_IS_LEAF)) + CCL_NAMESPACE_END #endif /* __KERNEL_TYPES_H__ */ diff --git a/intern/cycles/kernel/svm/svm_math_util.h b/intern/cycles/kernel/svm/svm_math_util.h index 3f7d18a02fe..6d13a0d8e02 100644 --- a/intern/cycles/kernel/svm/svm_math_util.h +++ b/intern/cycles/kernel/svm/svm_math_util.h @@ -32,21 +32,17 @@ ccl_device void svm_vector_math(float *Fac, float3 *Vector, NodeVectorMath type, *Fac = average_fac(*Vector); } else if(type == NODE_VECTOR_MATH_AVERAGE) { - *Fac = len(Vector1 + Vector2); - *Vector = normalize(Vector1 + Vector2); + *Vector = safe_normalize_len(Vector1 + Vector2, Fac); } else if(type == NODE_VECTOR_MATH_DOT_PRODUCT) { *Fac = dot(Vector1, Vector2); *Vector = make_float3(0.0f, 0.0f, 0.0f); } else if(type == NODE_VECTOR_MATH_CROSS_PRODUCT) { - float3 c = cross(Vector1, Vector2); - *Fac = len(c); - *Vector = normalize(c); + *Vector = safe_normalize_len(cross(Vector1, Vector2), Fac); } else if(type == NODE_VECTOR_MATH_NORMALIZE) { - *Fac = len(Vector1); - *Vector = normalize(Vector1); + *Vector = safe_normalize_len(Vector1, Fac); } else { *Fac = 0.0f; diff --git a/intern/cycles/render/attribute.cpp b/intern/cycles/render/attribute.cpp index a77ae1121a1..c0d429a583c 100644 --- a/intern/cycles/render/attribute.cpp +++ b/intern/cycles/render/attribute.cpp @@ -44,6 +44,7 @@ void Attribute::set(ustring name_, TypeDesc type_, AttributeElement element_) type = type_; element = element_; std = ATTR_STD_NONE; + flags = 0; /* string and matrix not supported! */ assert(type == TypeDesc::TypeFloat || type == TypeDesc::TypeColor || @@ -61,6 +62,11 @@ void Attribute::resize(Mesh *mesh, AttributePrimitive prim, bool reserve_only) } } +void Attribute::resize(size_t num_elements) +{ + buffer.resize(num_elements * data_sizeof(), 0); +} + void Attribute::add(const float& f) { char *data = (char*)&f; @@ -130,6 +136,10 @@ size_t Attribute::data_sizeof() const size_t Attribute::element_size(Mesh *mesh, AttributePrimitive prim) const { + if(flags & ATTR_FINAL_SIZE) { + return buffer.size() / data_sizeof(); + } + size_t size; switch(element) { diff --git a/intern/cycles/render/attribute.h b/intern/cycles/render/attribute.h index 67d067d60c9..f4538c76369 100644 --- a/intern/cycles/render/attribute.h +++ b/intern/cycles/render/attribute.h @@ -54,11 +54,13 @@ public: TypeDesc type; vector<char> buffer; AttributeElement element; + uint flags; /* enum AttributeFlag */ Attribute() {} ~Attribute(); void set(ustring name, TypeDesc type, AttributeElement element); void resize(Mesh *mesh, AttributePrimitive prim, bool reserve_only); + void resize(size_t num_elements); size_t data_sizeof() const; size_t element_size(Mesh *mesh, AttributePrimitive prim) const; diff --git a/intern/cycles/render/mesh.cpp b/intern/cycles/render/mesh.cpp index 2edec40b131..f90c19a11c8 100644 --- a/intern/cycles/render/mesh.cpp +++ b/intern/cycles/render/mesh.cpp @@ -30,6 +30,8 @@ #include "osl_globals.h" +#include "subd_patch_table.h" + #include "util_foreach.h" #include "util_logging.h" #include "util_progress.h" @@ -112,19 +114,12 @@ float3 Mesh::SubdFace::normal(const Mesh *mesh) const return safe_normalize(cross(v1 - v0, v2 - v0)); } - /* Mesh */ NODE_DEFINE(Mesh) { NodeType* type = NodeType::add("mesh", create); - static NodeEnum displacement_method_enum; - displacement_method_enum.insert("bump", DISPLACE_BUMP); - displacement_method_enum.insert("true", DISPLACE_TRUE); - displacement_method_enum.insert("both", DISPLACE_BOTH); - SOCKET_ENUM(displacement_method, "Displacement Method", displacement_method_enum, DISPLACE_BUMP); - SOCKET_UINT(motion_steps, "Motion Steps", 3); SOCKET_BOOLEAN(use_motion_blur, "Use Motion Blur", false); @@ -177,11 +172,14 @@ Mesh::Mesh() num_ngons = 0; subdivision_type = SUBDIVISION_NONE; + + patch_table = NULL; } Mesh::~Mesh() { delete bvh; + delete patch_table; } void Mesh::resize_mesh(int numverts, int numtris) @@ -274,6 +272,8 @@ void Mesh::clear() num_subd_verts = 0; + subd_creases.clear(); + attributes.clear(); curve_attributes.clear(); subd_attributes.clear(); @@ -283,6 +283,9 @@ void Mesh::clear() transform_negative_scaled = false; transform_normal = transform_identity(); geometry_flags = GEOMETRY_NONE; + + delete patch_table; + patch_table = NULL; } int Mesh::split_vertex(int vertex) @@ -705,7 +708,6 @@ void Mesh::pack_patches(uint *patch_data, uint vert_offset, uint face_offset, ui } } - void Mesh::compute_bvh(DeviceScene *dscene, SceneParams *params, Progress *progress, @@ -779,6 +781,17 @@ bool Mesh::has_motion_blur() const curve_attributes.find(ATTR_STD_MOTION_VERTEX_POSITION))); } +bool Mesh::has_true_displacement() const +{ + foreach(Shader *shader, used_shaders) { + if(shader->has_displacement && shader->displacement_method != DISPLACE_BUMP) { + return true; + } + } + + return false; +} + bool Mesh::need_build_bvh() const { return !transform_applied || has_surface_bssrdf; @@ -834,6 +847,7 @@ void MeshManager::update_osl_attributes(Device *device, Scene *scene, vector<Att osl_attr.desc.element = ATTR_ELEMENT_OBJECT; osl_attr.value = attr; osl_attr.desc.offset = 0; + osl_attr.desc.flags = 0; og->attribute_map[i*ATTR_PRIM_TYPES + ATTR_PRIM_TRIANGLE][attr.name()] = osl_attr; og->attribute_map[i*ATTR_PRIM_TYPES + ATTR_PRIM_CURVE][attr.name()] = osl_attr; @@ -977,6 +991,8 @@ void MeshManager::update_svm_attributes(Device *device, DeviceScene *dscene, Sce attr_map[index].w = NODE_ATTR_MATRIX; else attr_map[index].w = NODE_ATTR_FLOAT3; + + attr_map[index].w |= req.triangle_desc.flags << 8; } index++; @@ -992,6 +1008,8 @@ void MeshManager::update_svm_attributes(Device *device, DeviceScene *dscene, Sce attr_map[index].w = NODE_ATTR_MATRIX; else attr_map[index].w = NODE_ATTR_FLOAT3; + + attr_map[index].w |= req.curve_desc.flags << 8; } index++; @@ -1007,6 +1025,8 @@ void MeshManager::update_svm_attributes(Device *device, DeviceScene *dscene, Sce attr_map[index].w = NODE_ATTR_MATRIX; else attr_map[index].w = NODE_ATTR_FLOAT3; + + attr_map[index].w |= req.subd_desc.flags << 8; } index++; @@ -1071,6 +1091,7 @@ static void update_attribute_element_offset(Mesh *mesh, if(mattr) { /* store element and type */ desc.element = mattr->element; + desc.flags = mattr->flags; type = mattr->type; /* store attribute data in arrays */ @@ -1127,7 +1148,11 @@ static void update_attribute_element_offset(Mesh *mesh, /* mesh vertex/curve index is global, not per object, so we sneak * a correction for that in here */ - if(element == ATTR_ELEMENT_VERTEX) + if(mesh->subdivision_type == Mesh::SUBDIVISION_CATMULL_CLARK && desc.flags & ATTR_SUBDIVIDED) { + /* indices for subdivided attributes are retrieved + * from patch table so no need for correction here*/ + } + else if(element == ATTR_ELEMENT_VERTEX) offset -= mesh->vert_offset; else if(element == ATTR_ELEMENT_VERTEX_MOTION) offset -= mesh->vert_offset; @@ -1323,6 +1348,12 @@ void MeshManager::mesh_calc_offset(Scene *scene) if(mesh->subd_faces.size()) { Mesh::SubdFace& last = mesh->subd_faces[mesh->subd_faces.size()-1]; patch_size += (last.ptex_offset + last.num_ptex_faces()) * 8; + + /* patch tables are stored in same array so include them in patch_size */ + if(mesh->patch_table) { + mesh->patch_table_offset = patch_size; + patch_size += mesh->patch_table->total_size(); + } } face_size += mesh->subd_faces.size(); corner_size += mesh->subd_face_corners.size(); @@ -1354,6 +1385,12 @@ void MeshManager::device_update_mesh(Device *device, if(mesh->subd_faces.size()) { Mesh::SubdFace& last = mesh->subd_faces[mesh->subd_faces.size()-1]; patch_size += (last.ptex_offset + last.num_ptex_faces()) * 8; + + /* patch tables are stored in same array so include them in patch_size */ + if(mesh->patch_table) { + mesh->patch_table_offset = patch_size; + patch_size += mesh->patch_table->total_size(); + } } } @@ -1436,6 +1473,11 @@ void MeshManager::device_update_mesh(Device *device, foreach(Mesh *mesh, scene->meshes) { mesh->pack_patches(&patch_data[mesh->patch_offset], mesh->vert_offset, mesh->face_offset, mesh->corner_offset); + + if(mesh->patch_table) { + mesh->patch_table->copy_adjusting_offsets(&patch_data[mesh->patch_table_offset], mesh->patch_table_offset); + } + if(progress.get_cancel()) return; } @@ -1622,7 +1664,7 @@ void MeshManager::device_update(Device *device, DeviceScene *dscene, Scene *scen bool old_need_object_flags_update = false; foreach(Mesh *mesh, scene->meshes) { if(mesh->need_update && - mesh->displacement_method != Mesh::DISPLACE_BUMP) + mesh->has_true_displacement()) { true_displacement_used = true; break; @@ -1648,6 +1690,10 @@ void MeshManager::device_update(Device *device, DeviceScene *dscene, Scene *scen } if(progress.get_cancel()) return; + /* after mesh data has been copied to device memory we need to update + * offsets for patch tables as this can't be known before hand */ + scene->object_manager->device_update_patch_map_offsets(device, dscene, scene); + device_update_attributes(device, dscene, scene, progress); if(progress.get_cancel()) return; diff --git a/intern/cycles/render/mesh.h b/intern/cycles/render/mesh.h index c9ae9aab888..eff5c50e635 100644 --- a/intern/cycles/render/mesh.h +++ b/intern/cycles/render/mesh.h @@ -40,6 +40,7 @@ class Scene; class SceneParams; class AttributeRequest; class DiagSplit; +struct PackedPatchTable; /* Mesh */ @@ -110,13 +111,9 @@ public: int num_ptex_faces() const { return num_corners == 4 ? 1 : num_corners; } }; - /* Displacement */ - enum DisplacementMethod { - DISPLACE_BUMP = 0, - DISPLACE_TRUE = 1, - DISPLACE_BOTH = 2, - - DISPLACE_NUM_METHODS, + struct SubdEdgeCrease { + int v[2]; + float crease; }; enum SubdivisionType { @@ -157,6 +154,8 @@ public: array<int> subd_face_corners; int num_ngons; + array<SubdEdgeCrease> subd_creases; + vector<Shader*> used_shaders; AttributeSet attributes; AttributeSet curve_attributes; @@ -166,7 +165,8 @@ public: bool transform_applied; bool transform_negative_scaled; Transform transform_normal; - DisplacementMethod displacement_method; + + PackedPatchTable *patch_table; uint motion_steps; bool use_motion_blur; @@ -184,6 +184,7 @@ public: size_t curvekey_offset; size_t patch_offset; + size_t patch_table_offset; size_t face_offset; size_t corner_offset; @@ -234,6 +235,7 @@ public: void tag_update(Scene *scene, bool rebuild); bool has_motion_blur() const; + bool has_true_displacement() const; /* Check whether the mesh should have own BVH built separately. Briefly, * own BVH is needed for mesh, if: diff --git a/intern/cycles/render/mesh_displace.cpp b/intern/cycles/render/mesh_displace.cpp index 95f46ff02a2..ef9cfedd412 100644 --- a/intern/cycles/render/mesh_displace.cpp +++ b/intern/cycles/render/mesh_displace.cpp @@ -26,19 +26,27 @@ CCL_NAMESPACE_BEGIN +static float3 compute_face_normal(const Mesh::Triangle& t, float3 *verts) +{ + float3 v0 = verts[t.v[0]]; + float3 v1 = verts[t.v[1]]; + float3 v2 = verts[t.v[2]]; + + float3 norm = cross(v1 - v0, v2 - v0); + float normlen = len(norm); + + if(normlen == 0.0f) + return make_float3(1.0f, 0.0f, 0.0f); + + return norm / normlen; +} + bool MeshManager::displace(Device *device, DeviceScene *dscene, Scene *scene, Mesh *mesh, Progress& progress) { /* verify if we have a displacement shader */ - bool has_displacement = false; - - if(mesh->displacement_method != Mesh::DISPLACE_BUMP) { - foreach(Shader *shader, mesh->used_shaders) - if(shader->has_displacement) - has_displacement = true; - } - - if(!has_displacement) + if(!mesh->has_true_displacement()) { return false; + } string msg = string_printf("Computing Displacement %s", mesh->name.c_str()); progress.set_status("Updating Mesh", msg); @@ -67,8 +75,9 @@ bool MeshManager::displace(Device *device, DeviceScene *dscene, Scene *scene, Me Shader *shader = (shader_index < mesh->used_shaders.size()) ? mesh->used_shaders[shader_index] : scene->default_surface; - if(!shader->has_displacement) + if(!shader->has_displacement || shader->displacement_method == DISPLACE_BUMP) { continue; + } for(int j = 0; j < 3; j++) { if(done[t.v[j]]) @@ -153,8 +162,9 @@ bool MeshManager::displace(Device *device, DeviceScene *dscene, Scene *scene, Me Shader *shader = (shader_index < mesh->used_shaders.size()) ? mesh->used_shaders[shader_index] : scene->default_surface; - if(!shader->has_displacement) + if(!shader->has_displacement || shader->displacement_method == DISPLACE_BUMP) { continue; + } for(int j = 0; j < 3; j++) { if(!done[t.v[j]]) { @@ -178,9 +188,131 @@ bool MeshManager::displace(Device *device, DeviceScene *dscene, Scene *scene, Me mesh->attributes.remove(ATTR_STD_FACE_NORMAL); mesh->add_face_normals(); - if(mesh->displacement_method == Mesh::DISPLACE_TRUE) { - mesh->attributes.remove(ATTR_STD_VERTEX_NORMAL); - mesh->add_vertex_normals(); + bool need_recompute_vertex_normals = false; + + foreach(Shader *shader, mesh->used_shaders) { + if(shader->has_displacement && shader->displacement_method == DISPLACE_TRUE) { + need_recompute_vertex_normals = true; + break; + } + } + + if(need_recompute_vertex_normals) { + bool flip = mesh->transform_negative_scaled; + vector<bool> tri_has_true_disp(num_triangles, false); + + for(size_t i = 0; i < num_triangles; i++) { + int shader_index = mesh->shader[i]; + Shader *shader = (shader_index < mesh->used_shaders.size()) ? + mesh->used_shaders[shader_index] : scene->default_surface; + + tri_has_true_disp[i] = shader->has_displacement && shader->displacement_method == DISPLACE_TRUE; + } + + /* static vertex normals */ + + /* get attributes */ + Attribute *attr_fN = mesh->attributes.find(ATTR_STD_FACE_NORMAL); + Attribute *attr_vN = mesh->attributes.find(ATTR_STD_VERTEX_NORMAL); + + float3 *fN = attr_fN->data_float3(); + float3 *vN = attr_vN->data_float3(); + + /* compute vertex normals */ + + /* zero vertex normals on triangles with true displacement */ + for(size_t i = 0; i < num_triangles; i++) { + if(tri_has_true_disp[i]) { + for(size_t j = 0; j < 3; j++) { + vN[mesh->get_triangle(i).v[j]] = make_float3(0.0f, 0.0f, 0.0f); + } + } + } + + /* add face normals to vertex normals */ + for(size_t i = 0; i < num_triangles; i++) { + if(tri_has_true_disp[i]) { + for(size_t j = 0; j < 3; j++) { + vN[mesh->get_triangle(i).v[j]] += fN[i]; + } + } + } + + /* normalize vertex normals */ + done.clear(); + done.resize(num_verts, false); + + for(size_t i = 0; i < num_triangles; i++) { + if(tri_has_true_disp[i]) { + for(size_t j = 0; j < 3; j++) { + int vert = mesh->get_triangle(i).v[j]; + + if(done[vert]) { + continue; + } + + vN[vert] = normalize(vN[vert]); + if(flip) + vN[vert] = -vN[vert]; + + done[vert] = true; + } + } + } + + /* motion vertex normals */ + Attribute *attr_mP = mesh->attributes.find(ATTR_STD_MOTION_VERTEX_POSITION); + Attribute *attr_mN = mesh->attributes.find(ATTR_STD_MOTION_VERTEX_NORMAL); + + if(mesh->has_motion_blur() && attr_mP && attr_mN) { + for(int step = 0; step < mesh->motion_steps - 1; step++) { + float3 *mP = attr_mP->data_float3() + step*mesh->verts.size(); + float3 *mN = attr_mN->data_float3() + step*mesh->verts.size(); + + /* compute */ + + /* zero vertex normals on triangles with true displacement */ + for(size_t i = 0; i < num_triangles; i++) { + if(tri_has_true_disp[i]) { + for(size_t j = 0; j < 3; j++) { + mN[mesh->get_triangle(i).v[j]] = make_float3(0.0f, 0.0f, 0.0f); + } + } + } + + /* add face normals to vertex normals */ + for(size_t i = 0; i < num_triangles; i++) { + if(tri_has_true_disp[i]) { + for(size_t j = 0; j < 3; j++) { + float3 fN = compute_face_normal(mesh->get_triangle(i), mP); + mN[mesh->get_triangle(i).v[j]] += fN; + } + } + } + + /* normalize vertex normals */ + done.clear(); + done.resize(num_verts, false); + + for(size_t i = 0; i < num_triangles; i++) { + if(tri_has_true_disp[i]) { + for(size_t j = 0; j < 3; j++) { + int vert = mesh->get_triangle(i).v[j]; + + if(done[vert]) { + continue; + } + + mN[vert] = normalize(mN[vert]); + if(flip) + mN[vert] = -mN[vert]; + + done[vert] = true; + } + } + } + } + } } return true; diff --git a/intern/cycles/render/mesh_subdivision.cpp b/intern/cycles/render/mesh_subdivision.cpp index fe8e41e8d35..efb40efbb79 100644 --- a/intern/cycles/render/mesh_subdivision.cpp +++ b/intern/cycles/render/mesh_subdivision.cpp @@ -19,13 +19,302 @@ #include "subd_split.h" #include "subd_patch.h" +#include "subd_patch_table.h" #include "util_foreach.h" CCL_NAMESPACE_BEGIN +#ifdef WITH_OPENSUBDIV + +CCL_NAMESPACE_END + +#include <opensubdiv/far/topologyRefinerFactory.h> +#include <opensubdiv/far/primvarRefiner.h> +#include <opensubdiv/far/patchTableFactory.h> +#include <opensubdiv/far/patchMap.h> + +/* specializations of TopologyRefinerFactory for ccl::Mesh */ + +namespace OpenSubdiv { +namespace OPENSUBDIV_VERSION { +namespace Far { + template<> + bool TopologyRefinerFactory<ccl::Mesh>::resizeComponentTopology(TopologyRefiner& refiner, ccl::Mesh const& mesh) + { + setNumBaseVertices(refiner, mesh.verts.size()); + setNumBaseFaces(refiner, mesh.subd_faces.size()); + + ccl::Mesh::SubdFace* face = &mesh.subd_faces[0]; + + for(int i = 0; i < mesh.subd_faces.size(); i++, face++) { + setNumBaseFaceVertices(refiner, i, face->num_corners); + } + + return true; + } + + template<> + bool TopologyRefinerFactory<ccl::Mesh>::assignComponentTopology(TopologyRefiner& refiner, ccl::Mesh const& mesh) + { + ccl::Mesh::SubdFace* face = &mesh.subd_faces[0]; + + for(int i = 0; i < mesh.subd_faces.size(); i++, face++) { + IndexArray face_verts = getBaseFaceVertices(refiner, i); + + int* corner = &mesh.subd_face_corners[face->start_corner]; + + for(int j = 0; j < face->num_corners; j++, corner++) { + face_verts[j] = *corner; + } + } + + return true; + } + + template<> + bool TopologyRefinerFactory<ccl::Mesh>::assignComponentTags(TopologyRefiner& refiner, ccl::Mesh const& mesh) + { + const ccl::Mesh::SubdEdgeCrease* crease = mesh.subd_creases.data(); + + for(int i = 0; i < mesh.subd_creases.size(); i++, crease++) { + Index edge = findBaseEdge(refiner, crease->v[0], crease->v[1]); + + if(edge != INDEX_INVALID) { + setBaseEdgeSharpness(refiner, edge, crease->crease * 10.0f); + } + } + + for(int i = 0; i < mesh.verts.size(); i++) { + ConstIndexArray vert_edges = getBaseVertexEdges(refiner, i); + + if(vert_edges.size() == 2) { + float sharpness = refiner.getLevel(0).getEdgeSharpness(vert_edges[0]); + sharpness = std::min(sharpness, refiner.getLevel(0).getEdgeSharpness(vert_edges[1])); + + setBaseVertexSharpness(refiner, i, sharpness); + } + } + + return true; + } + + template<> + bool TopologyRefinerFactory<ccl::Mesh>::assignFaceVaryingTopology(TopologyRefiner& /*refiner*/, ccl::Mesh const& /*mesh*/) + { + return true; + } + + template<> + void TopologyRefinerFactory<ccl::Mesh>::reportInvalidTopology(TopologyError /*err_code*/, + char const */*msg*/, ccl::Mesh const& /*mesh*/) + { + } +} /* namespace Far */ +} /* namespace OPENSUBDIV_VERSION */ +} /* namespace OpenSubdiv */ + +CCL_NAMESPACE_BEGIN + +using namespace OpenSubdiv; + +/* struct that implements OpenSubdiv's vertex interface */ + +template<typename T> +struct OsdValue { + T value; + + OsdValue() {} + + void Clear(void* = 0) { + memset(&value, 0, sizeof(T)); + } + + void AddWithWeight(OsdValue<T> const& src, float weight) { + value += src.value * weight; + } +}; + +template<> +void OsdValue<uchar4>::AddWithWeight(OsdValue<uchar4> const& src, float weight) +{ + for(int i = 0; i < 4; i++) { + value[i] += (uchar)(src.value[i] * weight); + } +} + +/* class for holding OpenSubdiv data used during tessellation */ + +class OsdData { + Mesh* mesh; + vector<OsdValue<float3> > verts; + Far::TopologyRefiner* refiner; + Far::PatchTable* patch_table; + Far::PatchMap* patch_map; + +public: + OsdData() : mesh(NULL), refiner(NULL), patch_table(NULL), patch_map(NULL) {} + + ~OsdData() + { + delete refiner; + delete patch_table; + delete patch_map; + } + + void build_from_mesh(Mesh* mesh_) + { + mesh = mesh_; + + /* type and options */ + Sdc::SchemeType type = Sdc::SCHEME_CATMARK; + + Sdc::Options options; + options.SetVtxBoundaryInterpolation(Sdc::Options::VTX_BOUNDARY_EDGE_ONLY); + + /* create refiner */ + refiner = Far::TopologyRefinerFactory<Mesh>::Create(*mesh, + Far::TopologyRefinerFactory<Mesh>::Options(type, options)); + + /* adaptive refinement */ + int max_isolation = 10; + refiner->RefineAdaptive(Far::TopologyRefiner::AdaptiveOptions(max_isolation)); + + /* create patch table */ + Far::PatchTableFactory::Options patch_options; + patch_options.endCapType = Far::PatchTableFactory::Options::ENDCAP_GREGORY_BASIS; + + patch_table = Far::PatchTableFactory::Create(*refiner, patch_options); + + /* interpolate verts */ + int num_refiner_verts = refiner->GetNumVerticesTotal(); + int num_local_points = patch_table->GetNumLocalPoints(); + + verts.resize(num_refiner_verts + num_local_points); + for(int i = 0; i < mesh->verts.size(); i++) { + verts[i].value = mesh->verts[i]; + } + + OsdValue<float3>* src = &verts[0]; + for(int i = 0; i < refiner->GetMaxLevel(); i++) { + OsdValue<float3>* dest = src + refiner->GetLevel(i).GetNumVertices(); + Far::PrimvarRefiner(*refiner).Interpolate(i+1, src, dest); + src = dest; + } + + patch_table->ComputeLocalPointValues(&verts[0], &verts[num_refiner_verts]); + + /* create patch map */ + patch_map = new Far::PatchMap(*patch_table); + } + + void subdivide_attribute(Attribute& attr) + { + Far::PrimvarRefiner primvar_refiner(*refiner); + + if(attr.element == ATTR_ELEMENT_VERTEX) { + int num_refiner_verts = refiner->GetNumVerticesTotal(); + int num_local_points = patch_table->GetNumLocalPoints(); + + attr.resize(num_refiner_verts + num_local_points); + attr.flags |= ATTR_FINAL_SIZE; + + char* src = &attr.buffer[0]; + + for(int i = 0; i < refiner->GetMaxLevel(); i++) { + char* dest = src + refiner->GetLevel(i).GetNumVertices() * attr.data_sizeof(); + + if(attr.same_storage(attr.type, TypeDesc::TypeFloat)) { + primvar_refiner.Interpolate(i+1, (OsdValue<float>*)src, (OsdValue<float>*&)dest); + } + else { + primvar_refiner.Interpolate(i+1, (OsdValue<float4>*)src, (OsdValue<float4>*&)dest); + } + + src = dest; + } + + if(attr.same_storage(attr.type, TypeDesc::TypeFloat)) { + patch_table->ComputeLocalPointValues((OsdValue<float>*)&attr.buffer[0], + (OsdValue<float>*)&attr.buffer[num_refiner_verts * attr.data_sizeof()]); + } + else { + patch_table->ComputeLocalPointValues((OsdValue<float4>*)&attr.buffer[0], + (OsdValue<float4>*)&attr.buffer[num_refiner_verts * attr.data_sizeof()]); + } + } + else if(attr.element == ATTR_ELEMENT_CORNER || attr.element == ATTR_ELEMENT_CORNER_BYTE) { + // TODO(mai): fvar interpolation + } + } + + friend struct OsdPatch; + friend class Mesh; +}; + +/* ccl::Patch implementation that uses OpenSubdiv for eval */ + +struct OsdPatch : Patch { + OsdData* osd_data; + + OsdPatch(OsdData* data) : osd_data(data) {} + + void eval(float3 *P, float3 *dPdu, float3 *dPdv, float3 *N, float u, float v) + { + const Far::PatchTable::PatchHandle* handle = osd_data->patch_map->FindPatch(patch_index, u, v); + assert(handle); + + float p_weights[20], du_weights[20], dv_weights[20]; + osd_data->patch_table->EvaluateBasis(*handle, u, v, p_weights, du_weights, dv_weights); + + Far::ConstIndexArray cv = osd_data->patch_table->GetPatchVertices(*handle); + + float3 du, dv; + if(P) *P = make_float3(0.0f, 0.0f, 0.0f); + du = make_float3(0.0f, 0.0f, 0.0f); + dv = make_float3(0.0f, 0.0f, 0.0f); + + for(int i = 0; i < cv.size(); i++) { + float3 p = osd_data->verts[cv[i]].value; + + if(P) *P += p * p_weights[i]; + du += p * du_weights[i]; + dv += p * dv_weights[i]; + } + + if(dPdu) *dPdu = du; + if(dPdv) *dPdv = dv; + if(N) *N = normalize(cross(du, dv)); + } + + BoundBox bound() { return BoundBox::empty; } +}; + +#endif + void Mesh::tessellate(DiagSplit *split) { +#ifdef WITH_OPENSUBDIV + OsdData osd_data; + bool need_packed_patch_table = false; + + if(subdivision_type == SUBDIVISION_CATMULL_CLARK) { + osd_data.build_from_mesh(this); + } + else +#endif + { + /* force linear subdivision if OpenSubdiv is unavailable to avoid + * falling into catmull-clark code paths by accident + */ + subdivision_type = SUBDIVISION_LINEAR; + + /* force disable attribute subdivision for same reason as above */ + foreach(Attribute& attr, subd_attributes.attributes) { + attr.flags &= ~ATTR_SUBDIVIDED; + } + } + int num_faces = subd_faces.size(); Attribute *attr_vN = subd_attributes.find(ATTR_STD_VERTEX_NORMAL); @@ -36,113 +325,158 @@ void Mesh::tessellate(DiagSplit *split) if(face.is_quad()) { /* quad */ - LinearQuadPatch patch; - float3 *hull = patch.hull; - float3 *normals = patch.normals; + QuadDice::SubPatch subpatch; - patch.patch_index = face.ptex_offset; - patch.shader = face.shader; + LinearQuadPatch quad_patch; +#ifdef WITH_OPENSUBDIV + OsdPatch osd_patch(&osd_data); - for(int i = 0; i < 4; i++) { - hull[i] = verts[subd_face_corners[face.start_corner+i]]; + if(subdivision_type == SUBDIVISION_CATMULL_CLARK) { + osd_patch.patch_index = face.ptex_offset; + + subpatch.patch = &osd_patch; } + else +#endif + { + float3 *hull = quad_patch.hull; + float3 *normals = quad_patch.normals; + + quad_patch.patch_index = face.ptex_offset; - if(face.smooth) { for(int i = 0; i < 4; i++) { - normals[i] = vN[subd_face_corners[face.start_corner+i]]; + hull[i] = verts[subd_face_corners[face.start_corner+i]]; } - } - else { - float3 N = face.normal(this); - for(int i = 0; i < 4; i++) { - normals[i] = N; + + if(face.smooth) { + for(int i = 0; i < 4; i++) { + normals[i] = vN[subd_face_corners[face.start_corner+i]]; + } + } + else { + float3 N = face.normal(this); + for(int i = 0; i < 4; i++) { + normals[i] = N; + } } + + swap(hull[2], hull[3]); + swap(normals[2], normals[3]); + + subpatch.patch = &quad_patch; } - swap(hull[2], hull[3]); - swap(normals[2], normals[3]); + subpatch.patch->shader = face.shader; /* Quad faces need to be split at least once to line up with split ngons, we do this * here in this manner because if we do it later edge factors may end up slightly off. */ - QuadDice::SubPatch subpatch; - subpatch.patch = &patch; - subpatch.P00 = make_float2(0.0f, 0.0f); subpatch.P10 = make_float2(0.5f, 0.0f); subpatch.P01 = make_float2(0.0f, 0.5f); subpatch.P11 = make_float2(0.5f, 0.5f); - split->split_quad(&patch, &subpatch); + split->split_quad(subpatch.patch, &subpatch); subpatch.P00 = make_float2(0.5f, 0.0f); subpatch.P10 = make_float2(1.0f, 0.0f); subpatch.P01 = make_float2(0.5f, 0.5f); subpatch.P11 = make_float2(1.0f, 0.5f); - split->split_quad(&patch, &subpatch); + split->split_quad(subpatch.patch, &subpatch); subpatch.P00 = make_float2(0.0f, 0.5f); subpatch.P10 = make_float2(0.5f, 0.5f); subpatch.P01 = make_float2(0.0f, 1.0f); subpatch.P11 = make_float2(0.5f, 1.0f); - split->split_quad(&patch, &subpatch); + split->split_quad(subpatch.patch, &subpatch); subpatch.P00 = make_float2(0.5f, 0.5f); subpatch.P10 = make_float2(1.0f, 0.5f); subpatch.P01 = make_float2(0.5f, 1.0f); subpatch.P11 = make_float2(1.0f, 1.0f); - split->split_quad(&patch, &subpatch); + split->split_quad(subpatch.patch, &subpatch); } else { /* ngon */ - float3 center_vert = make_float3(0.0f, 0.0f, 0.0f); - float3 center_normal = make_float3(0.0f, 0.0f, 0.0f); +#ifdef WITH_OPENSUBDIV + if(subdivision_type == SUBDIVISION_CATMULL_CLARK) { + OsdPatch patch(&osd_data); + + patch.shader = face.shader; - float inv_num_corners = 1.0f/float(face.num_corners); - for(int corner = 0; corner < face.num_corners; corner++) { - center_vert += verts[subd_face_corners[face.start_corner + corner]] * inv_num_corners; - center_normal += vN[subd_face_corners[face.start_corner + corner]] * inv_num_corners; + for(int corner = 0; corner < face.num_corners; corner++) { + patch.patch_index = face.ptex_offset + corner; + + split->split_quad(&patch); + } } + else +#endif + { + float3 center_vert = make_float3(0.0f, 0.0f, 0.0f); + float3 center_normal = make_float3(0.0f, 0.0f, 0.0f); + + float inv_num_corners = 1.0f/float(face.num_corners); + for(int corner = 0; corner < face.num_corners; corner++) { + center_vert += verts[subd_face_corners[face.start_corner + corner]] * inv_num_corners; + center_normal += vN[subd_face_corners[face.start_corner + corner]] * inv_num_corners; + } - for(int corner = 0; corner < face.num_corners; corner++) { - LinearQuadPatch patch; - float3 *hull = patch.hull; - float3 *normals = patch.normals; + for(int corner = 0; corner < face.num_corners; corner++) { + LinearQuadPatch patch; + float3 *hull = patch.hull; + float3 *normals = patch.normals; - patch.patch_index = face.ptex_offset + corner; + patch.patch_index = face.ptex_offset + corner; - patch.shader = face.shader; + patch.shader = face.shader; - hull[0] = verts[subd_face_corners[face.start_corner + mod(corner + 0, face.num_corners)]]; - hull[1] = verts[subd_face_corners[face.start_corner + mod(corner + 1, face.num_corners)]]; - hull[2] = verts[subd_face_corners[face.start_corner + mod(corner - 1, face.num_corners)]]; - hull[3] = center_vert; + hull[0] = verts[subd_face_corners[face.start_corner + mod(corner + 0, face.num_corners)]]; + hull[1] = verts[subd_face_corners[face.start_corner + mod(corner + 1, face.num_corners)]]; + hull[2] = verts[subd_face_corners[face.start_corner + mod(corner - 1, face.num_corners)]]; + hull[3] = center_vert; - hull[1] = (hull[1] + hull[0]) * 0.5; - hull[2] = (hull[2] + hull[0]) * 0.5; + hull[1] = (hull[1] + hull[0]) * 0.5; + hull[2] = (hull[2] + hull[0]) * 0.5; - if(face.smooth) { - normals[0] = vN[subd_face_corners[face.start_corner + mod(corner + 0, face.num_corners)]]; - normals[1] = vN[subd_face_corners[face.start_corner + mod(corner + 1, face.num_corners)]]; - normals[2] = vN[subd_face_corners[face.start_corner + mod(corner - 1, face.num_corners)]]; - normals[3] = center_normal; + if(face.smooth) { + normals[0] = vN[subd_face_corners[face.start_corner + mod(corner + 0, face.num_corners)]]; + normals[1] = vN[subd_face_corners[face.start_corner + mod(corner + 1, face.num_corners)]]; + normals[2] = vN[subd_face_corners[face.start_corner + mod(corner - 1, face.num_corners)]]; + normals[3] = center_normal; - normals[1] = (normals[1] + normals[0]) * 0.5; - normals[2] = (normals[2] + normals[0]) * 0.5; - } - else { - float3 N = face.normal(this); - for(int i = 0; i < 4; i++) { - normals[i] = N; + normals[1] = (normals[1] + normals[0]) * 0.5; + normals[2] = (normals[2] + normals[0]) * 0.5; + } + else { + float3 N = face.normal(this); + for(int i = 0; i < 4; i++) { + normals[i] = N; + } } - } - split->split_quad(&patch); + split->split_quad(&patch); + } } } } /* interpolate center points for attributes */ foreach(Attribute& attr, subd_attributes.attributes) { +#ifdef WITH_OPENSUBDIV + if(subdivision_type == SUBDIVISION_CATMULL_CLARK && attr.flags & ATTR_SUBDIVIDED) { + if(attr.element == ATTR_ELEMENT_CORNER || attr.element == ATTR_ELEMENT_CORNER_BYTE) { + /* keep subdivision for corner attributes disabled for now */ + attr.flags &= ~ATTR_SUBDIVIDED; + } + else { + osd_data.subdivide_attribute(attr); + + need_packed_patch_table = true; + continue; + } + } +#endif + char* data = attr.data(); size_t stride = attr.data_sizeof(); int ngons = 0; @@ -218,6 +552,15 @@ void Mesh::tessellate(DiagSplit *split) default: break; } } + +#ifdef WITH_OPENSUBDIV + /* pack patch tables */ + if(need_packed_patch_table) { + delete patch_table; + patch_table = new PackedPatchTable; + patch_table->pack(osd_data.patch_table); + } +#endif } CCL_NAMESPACE_END diff --git a/intern/cycles/render/object.cpp b/intern/cycles/render/object.cpp index 5dbe2e4d6a1..62076f3a865 100644 --- a/intern/cycles/render/object.cpp +++ b/intern/cycles/render/object.cpp @@ -29,6 +29,8 @@ #include "util_progress.h" #include "util_vector.h" +#include "subd_patch_table.h" + CCL_NAMESPACE_BEGIN /* Object */ @@ -607,6 +609,40 @@ void ObjectManager::device_update_flags(Device *device, device->tex_alloc("__object_flag", dscene->object_flag); } +void ObjectManager::device_update_patch_map_offsets(Device *device, DeviceScene *dscene, Scene *scene) +{ + if (scene->objects.size() == 0) + return; + + uint4* objects = (uint4*)dscene->objects.get_data(); + + bool update = false; + + int object_index = 0; + foreach(Object *object, scene->objects) { + int offset = object_index*OBJECT_SIZE + 11; + + Mesh* mesh = object->mesh; + + if(mesh->patch_table) { + uint patch_map_offset = 2*(mesh->patch_table_offset + mesh->patch_table->total_size() - + mesh->patch_table->num_nodes * PATCH_NODE_SIZE) - mesh->patch_offset; + + if(objects[offset].x != patch_map_offset) { + objects[offset].x = patch_map_offset; + update = true; + } + } + + object_index++; + } + + if(update) { + device->tex_free(dscene->objects); + device->tex_alloc("__objects", dscene->objects); + } +} + void ObjectManager::device_free(Device *device, DeviceScene *dscene) { device->tex_free(dscene->objects); @@ -656,7 +692,7 @@ void ObjectManager::apply_static_transforms(DeviceScene *dscene, Scene *scene, u * Could be solved by moving reference counter to Mesh. */ if((mesh_users[object->mesh] == 1 && !object->mesh->has_surface_bssrdf) && - object->mesh->displacement_method == Mesh::DISPLACE_BUMP) + !object->mesh->has_true_displacement()) { if(!(motion_blur && object->use_motion)) { if(!object->mesh->transform_applied) { diff --git a/intern/cycles/render/object.h b/intern/cycles/render/object.h index 7ab73f3c91a..2e5837f672f 100644 --- a/intern/cycles/render/object.h +++ b/intern/cycles/render/object.h @@ -97,6 +97,8 @@ public: Scene *scene, Progress& progress, bool bounds_valid = true); + void device_update_patch_map_offsets(Device *device, DeviceScene *dscene, Scene *scene); + void device_free(Device *device, DeviceScene *dscene); void tag_update(Scene *scene); diff --git a/intern/cycles/render/shader.cpp b/intern/cycles/render/shader.cpp index 4cdb878df45..d000cca5a45 100644 --- a/intern/cycles/render/shader.cpp +++ b/intern/cycles/render/shader.cpp @@ -150,6 +150,12 @@ NODE_DEFINE(Shader) volume_interpolation_method_enum.insert("cubic", VOLUME_INTERPOLATION_CUBIC); SOCKET_ENUM(volume_interpolation_method, "Volume Interpolation Method", volume_interpolation_method_enum, VOLUME_INTERPOLATION_LINEAR); + static NodeEnum displacement_method_enum; + displacement_method_enum.insert("bump", DISPLACE_BUMP); + displacement_method_enum.insert("true", DISPLACE_TRUE); + displacement_method_enum.insert("both", DISPLACE_BOTH); + SOCKET_ENUM(displacement_method, "Displacement Method", displacement_method_enum, DISPLACE_BUMP); + return type; } @@ -173,6 +179,8 @@ Shader::Shader() has_object_dependency = false; has_integrator_dependency = false; + displacement_method = DISPLACE_BUMP; + id = -1; used = false; @@ -310,7 +318,7 @@ int ShaderManager::get_shader_id(Shader *shader, Mesh *mesh, bool smooth) int id = shader->id*2; /* index depends bump since this setting is not in the shader */ - if(mesh && mesh->displacement_method != Mesh::DISPLACE_TRUE) + if(mesh && shader->displacement_method != DISPLACE_TRUE) id += 1; /* smooth flag */ if(smooth) diff --git a/intern/cycles/render/shader.h b/intern/cycles/render/shader.h index dc57ed4e4eb..060ad7056bc 100644 --- a/intern/cycles/render/shader.h +++ b/intern/cycles/render/shader.h @@ -66,6 +66,14 @@ enum VolumeInterpolation { VOLUME_NUM_INTERPOLATION, }; +enum DisplacementMethod { + DISPLACE_BUMP = 0, + DISPLACE_TRUE = 1, + DISPLACE_BOTH = 2, + + DISPLACE_NUM_METHODS, +}; + /* Shader describing the appearance of a Mesh, Light or Background. * * While there is only a single shader graph, it has three outputs: surface, @@ -110,6 +118,9 @@ public: bool has_object_dependency; bool has_integrator_dependency; + /* displacement */ + DisplacementMethod displacement_method; + /* requested mesh attributes */ AttributeRequestSet attributes; diff --git a/intern/cycles/subd/CMakeLists.txt b/intern/cycles/subd/CMakeLists.txt index db497013693..9265299e82b 100644 --- a/intern/cycles/subd/CMakeLists.txt +++ b/intern/cycles/subd/CMakeLists.txt @@ -16,6 +16,7 @@ set(SRC subd_dice.cpp subd_patch.cpp subd_split.cpp + subd_patch_table.cpp ) set(SRC_HEADERS @@ -24,10 +25,6 @@ set(SRC_HEADERS subd_split.h ) -if(WITH_CYCLES_OPENSUBDIV) - add_definitions(-DWITH_OPENSUBDIV) -endif() - include_directories(${INC}) include_directories(SYSTEM ${INC_SYS}) diff --git a/intern/cycles/subd/subd_patch_table.cpp b/intern/cycles/subd/subd_patch_table.cpp new file mode 100644 index 00000000000..68ec1b2c6a6 --- /dev/null +++ b/intern/cycles/subd/subd_patch_table.cpp @@ -0,0 +1,297 @@ +/* + * Based on code from OpenSubdiv released under this license: + * + * Copyright 2014 DreamWorks Animation LLC. + * + * Licensed under the Apache License, Version 2.0 (the "Apache License") + * with the following modification; you may not use this file except in + * compliance with the Apache License and the following modification to it: + * Section 6. Trademarks. is deleted and replaced with: + * + * 6. Trademarks. This License does not grant permission to use the trade + * names, trademarks, service marks, or product names of the Licensor + * and its affiliates, except as required to comply with Section 4(c) of + * the License and to reproduce the content of the NOTICE file. + * + * You may obtain a copy of the Apache License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the Apache License with the above modification is + * distributed on an "AS IS" BASIS, WITHOUT WARRANTIES OR CONDITIONS OF ANY + * KIND, either express or implied. See the Apache License for the specific + * language governing permissions and limitations under the Apache License. + * + */ + +#include "subd_patch_table.h" +#include "kernel_types.h" + +#include "util_math.h" + +#ifdef WITH_OPENSUBDIV +#include <opensubdiv/far/patchTable.h> +#endif + +CCL_NAMESPACE_BEGIN + +#ifdef WITH_OPENSUBDIV + +using namespace OpenSubdiv; + +/* functions for building patch maps */ + +struct PatchMapQuadNode { + /* sets all the children to point to the patch of index */ + void set_child(int index) + { + for (int i = 0; i < 4; i++) { + children[i] = index | PATCH_MAP_NODE_IS_SET | PATCH_MAP_NODE_IS_LEAF; + } + } + + /* sets the child in quadrant to point to the node or patch of the given index */ + void set_child(unsigned char quadrant, int index, bool is_leaf=true) + { + assert(quadrant < 4); + children[quadrant] = index | PATCH_MAP_NODE_IS_SET | (is_leaf ? PATCH_MAP_NODE_IS_LEAF : 0); + } + + uint children[4]; +}; + +template<class T> +static int resolve_quadrant(T& median, T& u, T& v) +{ + int quadrant = -1; + + if(u < median) { + if(v < median) { + quadrant = 0; + } + else { + quadrant = 1; + v -= median; + } + } + else { + if(v < median) { + quadrant = 3; + } + else { + quadrant = 2; + v -= median; + } + u -= median; + } + + return quadrant; +} + +static void build_patch_map(PackedPatchTable& table, OpenSubdiv::Far::PatchTable* patch_table, int offset) +{ + int num_faces = 0; + + for(int array = 0; array < table.num_arrays; array++) { + Far::ConstPatchParamArray params = patch_table->GetPatchParams(array); + + for(int j = 0; j < patch_table->GetNumPatches(array); j++) { + num_faces = max(num_faces, (int)params[j].GetFaceId()); + } + } + num_faces++; + + vector<PatchMapQuadNode> quadtree; + quadtree.reserve(num_faces + table.num_patches); + quadtree.resize(num_faces); + + /* adjust offsets to make indices relative to the table */ + int handle_index = -(table.num_patches * PATCH_HANDLE_SIZE); + offset += table.total_size(); + + /* populate the quadtree from the FarPatchArrays sub-patches */ + for(int array = 0; array < table.num_arrays; array++) { + Far::ConstPatchParamArray params = patch_table->GetPatchParams(array); + + for(int i = 0; i < patch_table->GetNumPatches(array); i++, handle_index += PATCH_HANDLE_SIZE) { + const Far::PatchParam& param = params[i]; + unsigned short depth = param.GetDepth(); + + PatchMapQuadNode* node = &quadtree[params[i].GetFaceId()]; + + if(depth == (param.NonQuadRoot() ? 1 : 0)) { + /* special case : regular BSpline face w/ no sub-patches */ + node->set_child(handle_index + offset); + continue; + } + + int u = param.GetU(); + int v = param.GetV(); + int pdepth = param.NonQuadRoot() ? depth-2 : depth-1; + int half = 1 << pdepth; + + for(int j = 0; j < depth; j++) { + int delta = half >> 1; + + int quadrant = resolve_quadrant(half, u, v); + assert(quadrant >= 0); + + half = delta; + + if(j == pdepth) { + /* we have reached the depth of the sub-patch : add a leaf */ + assert(!(node->children[quadrant] & PATCH_MAP_NODE_IS_SET)); + node->set_child(quadrant, handle_index + offset, true); + break; + } + else { + /* travel down the child node of the corresponding quadrant */ + if(!(node->children[quadrant] & PATCH_MAP_NODE_IS_SET)) { + /* create a new branch in the quadrant */ + quadtree.push_back(PatchMapQuadNode()); + + int idx = (int)quadtree.size() - 1; + node->set_child(quadrant, idx*4 + offset, false); + + node = &quadtree[idx]; + } + else { + /* travel down an existing branch */ + uint idx = node->children[quadrant] & PATCH_MAP_NODE_INDEX_MASK; + node = &(quadtree[(idx - offset)/4]); + } + } + } + } + } + + /* copy into table */ + assert(table.table.size() == table.total_size()); + uint map_offset = table.total_size(); + + table.num_nodes = quadtree.size() * 4; + table.table.resize(table.total_size()); + + uint* data = &table.table[map_offset]; + + for(int i = 0; i < quadtree.size(); i++) { + for(int j = 0; j < 4; j++) { + assert(quadtree[i].children[j] & PATCH_MAP_NODE_IS_SET); + *(data++) = quadtree[i].children[j]; + } + } +} + +#endif + +/* packed patch table functions */ + +size_t PackedPatchTable::total_size() +{ + return num_arrays * PATCH_ARRAY_SIZE + + num_indices + + num_patches * (PATCH_PARAM_SIZE + PATCH_HANDLE_SIZE) + + num_nodes * PATCH_NODE_SIZE; +} + +void PackedPatchTable::pack(Far::PatchTable* patch_table, int offset) +{ + num_arrays = 0; + num_patches = 0; + num_indices = 0; + num_nodes = 0; + +#ifdef WITH_OPENSUBDIV + num_arrays = patch_table->GetNumPatchArrays(); + + for(int i = 0; i < num_arrays; i++) { + int patches = patch_table->GetNumPatches(i); + int num_control = patch_table->GetPatchArrayDescriptor(i).GetNumControlVertices(); + + num_patches += patches; + num_indices += patches * num_control; + } + + table.resize(total_size()); + uint* data = &table[0]; + + uint* array = data; + uint* index = array + num_arrays * PATCH_ARRAY_SIZE; + uint* param = index + num_indices; + uint* handle = param + num_patches * PATCH_PARAM_SIZE; + + uint current_param = 0; + + for(int i = 0; i < num_arrays; i++) { + *(array++) = patch_table->GetPatchArrayDescriptor(i).GetType(); + *(array++) = patch_table->GetNumPatches(i); + *(array++) = (index - data) + offset; + *(array++) = (param - data) + offset; + + Far::ConstIndexArray indices = patch_table->GetPatchArrayVertices(i); + + for(int j = 0; j < indices.size(); j++) { + *(index++) = indices[j]; + } + + const Far::PatchParamTable& param_table = patch_table->GetPatchParamTable(); + + int num_control = patch_table->GetPatchArrayDescriptor(i).GetNumControlVertices(); + int patches = patch_table->GetNumPatches(i); + + for(int j = 0; j < patches; j++, current_param++) { + *(param++) = param_table[current_param].field0; + *(param++) = param_table[current_param].field1; + + *(handle++) = (array - data) - PATCH_ARRAY_SIZE + offset; + *(handle++) = (param - data) - PATCH_PARAM_SIZE + offset; + *(handle++) = j * num_control; + } + } + + build_patch_map(*this, patch_table, offset); +#else + (void)patch_table; + (void)offset; +#endif +} + +void PackedPatchTable::copy_adjusting_offsets(uint* dest, int doffset) +{ + uint* src = &table[0]; + + /* arrays */ + for(int i = 0; i < num_arrays; i++) { + *(dest++) = *(src++); + *(dest++) = *(src++); + *(dest++) = *(src++) + doffset; + *(dest++) = *(src++) + doffset; + } + + /* indices */ + for(int i = 0; i < num_indices; i++) { + *(dest++) = *(src++); + } + + /* params */ + for(int i = 0; i < num_patches; i++) { + *(dest++) = *(src++); + *(dest++) = *(src++); + } + + /* handles */ + for(int i = 0; i < num_patches; i++) { + *(dest++) = *(src++) + doffset; + *(dest++) = *(src++) + doffset; + *(dest++) = *(src++); + } + + /* nodes */ + for(int i = 0; i < num_nodes; i++) { + *(dest++) = *(src++) + doffset; + } +} + +CCL_NAMESPACE_END + diff --git a/intern/cycles/subd/subd_patch_table.h b/intern/cycles/subd/subd_patch_table.h new file mode 100644 index 00000000000..c8c7ecf9e47 --- /dev/null +++ b/intern/cycles/subd/subd_patch_table.h @@ -0,0 +1,63 @@ +/* + * Copyright 2011-2016 Blender Foundation + * + * Licensed under the Apache License, Version 2.0 (the "License"); + * you may not use this file except in compliance with the License. + * You may obtain a copy of the License at + * + * http://www.apache.org/licenses/LICENSE-2.0 + * + * Unless required by applicable law or agreed to in writing, software + * distributed under the License is distributed on an "AS IS" BASIS, + * WITHOUT WARRANTIES OR CONDITIONS OF ANY KIND, either express or implied. + * See the License for the specific language governing permissions and + * limitations under the License. + */ + +#ifndef __SUBD_PATCH_TABLE_H__ +#define __SUBD_PATCH_TABLE_H__ + +#include "util_types.h" +#include "util_vector.h" + +#ifdef WITH_OPENSUBDIV +#ifdef _MSC_VER +# include "iso646.h" +#endif + +#include <opensubdiv/far/patchTable.h> +#endif + +CCL_NAMESPACE_BEGIN + +#ifdef WITH_OPENSUBDIV +using namespace OpenSubdiv; +#else +/* forward declare for when OpenSubdiv is unavailable */ +namespace Far { struct PatchTable; } +#endif + +#define PATCH_ARRAY_SIZE 4 +#define PATCH_PARAM_SIZE 2 +#define PATCH_HANDLE_SIZE 3 +#define PATCH_NODE_SIZE 1 + +struct PackedPatchTable { + vector<uint> table; + + size_t num_arrays; + size_t num_indices; + size_t num_patches; + size_t num_nodes; + + /* calculated size from num_* members */ + size_t total_size(); + + void pack(Far::PatchTable* patch_table, int offset = 0); + void copy_adjusting_offsets(uint* dest, int doffset); +}; + +CCL_NAMESPACE_END + +#endif /* __SUBD_PATCH_TABLE_H__ */ + diff --git a/intern/cycles/test/CMakeLists.txt b/intern/cycles/test/CMakeLists.txt index 80fe893826a..9af777fb9dd 100644 --- a/intern/cycles/test/CMakeLists.txt +++ b/intern/cycles/test/CMakeLists.txt @@ -26,6 +26,7 @@ set(ALL_CYCLES_LIBRARIES cycles_device cycles_bvh cycles_graph + cycles_subd cycles_util ${OPENIMAGEIO_LIBRARIES} ) @@ -41,6 +42,16 @@ if(WITH_IMAGE_OPENJPEG AND NOT WITH_SYSTEM_OPENJPEG) extern_openjpeg ) endif() +if(WITH_CYCLES_OPENSUBDIV) + add_definitions(-DWITH_OPENSUBDIV) + include_directories( + SYSTEM + ${OPENSUBDIV_INCLUDE_DIR} + ) + list(APPEND ALL_CYCLES_LIBRARIES + ${OPENSUBDIV_LIBRARIES} + ) +endif() list(APPEND ALL_CYCLES_LIBRARIES ${BOOST_LIBRARIES} ) diff --git a/intern/cycles/util/util_math.h b/intern/cycles/util/util_math.h index 13aba0646d2..89a882d9b9d 100644 --- a/intern/cycles/util/util_math.h +++ b/intern/cycles/util/util_math.h @@ -572,6 +572,12 @@ ccl_device_inline float3 safe_normalize(const float3 a) return (t != 0.0f)? a/t: a; } +ccl_device_inline float3 safe_normalize_len(const float3 a, float *t) +{ + *t = len(a); + return (*t != 0.0f)? a/(*t): a; +} + #ifndef __KERNEL_OPENCL__ ccl_device_inline bool operator==(const float3 a, const float3 b) diff --git a/intern/cycles/util/util_vector.h b/intern/cycles/util/util_vector.h index 6f8c3f6f3de..546b17570bb 100644 --- a/intern/cycles/util/util_vector.h +++ b/intern/cycles/util/util_vector.h @@ -222,6 +222,11 @@ public: return datasize_; } + T* data() + { + return data_; + } + const T* data() const { return data_; diff --git a/release/scripts/startup/bl_ui/properties_grease_pencil_common.py b/release/scripts/startup/bl_ui/properties_grease_pencil_common.py index c80c5ca7ddd..90cf410036a 100644 --- a/release/scripts/startup/bl_ui/properties_grease_pencil_common.py +++ b/release/scripts/startup/bl_ui/properties_grease_pencil_common.py @@ -80,6 +80,7 @@ class GreasePencilDrawingToolsPanel: sub = col.column(align=True) sub.prop(context.tool_settings, "use_gpencil_additive_drawing", text="Additive Drawing") sub.prop(context.tool_settings, "use_gpencil_continuous_drawing", text="Continuous Drawing") + sub.prop(context.tool_settings, "use_gpencil_draw_onback", text="Draw on Back") col.separator() col.separator() @@ -658,7 +659,7 @@ class GPENCIL_UL_palettecolor(UIList): row = split.row(align=True) row.prop(palcolor, "color", text="", emboss=palcolor.is_stroke_visible) row.prop(palcolor, "fill_color", text="", emboss=palcolor.is_fill_visible) - split.prop(palcolor, "info", text="", emboss=False) + split.prop(palcolor, "name", text="", emboss=False) row = layout.row(align=True) row.prop(palcolor, "lock", text="", emboss=False) diff --git a/source/blender/alembic/CMakeLists.txt b/source/blender/alembic/CMakeLists.txt index f579a3b200a..42bd6a9c340 100644 --- a/source/blender/alembic/CMakeLists.txt +++ b/source/blender/alembic/CMakeLists.txt @@ -40,7 +40,7 @@ set(INC_SYS ${HDF5_INCLUDE_DIRS} ${OPENEXR_INCLUDE_DIRS} ) -if(APPLE) +if(APPLE OR WIN32) list(APPEND INC_SYS ${BOOST_INCLUDE_DIR} ) diff --git a/source/blender/blenkernel/BKE_idcode.h b/source/blender/blenkernel/BKE_idcode.h index 6de0efe2709..964a49435f1 100644 --- a/source/blender/blenkernel/BKE_idcode.h +++ b/source/blender/blenkernel/BKE_idcode.h @@ -42,6 +42,8 @@ bool BKE_idcode_is_valid(short idcode); int BKE_idcode_to_idfilter(const short idcode); short BKE_idcode_from_idfilter(const int idfilter); +int BKE_idcode_to_index(const short idcode); + /** * Return an ID code and steps the index forward 1. * diff --git a/source/blender/blenkernel/BKE_main.h b/source/blender/blenkernel/BKE_main.h index 6487954844b..7eba01e6d5a 100644 --- a/source/blender/blenkernel/BKE_main.h +++ b/source/blender/blenkernel/BKE_main.h @@ -42,6 +42,8 @@ */ #include "DNA_listBase.h" +#include "BKE_library.h" + #ifdef __cplusplus extern "C" { #endif @@ -103,7 +105,7 @@ typedef struct Main { ListBase linestyle; ListBase cachefiles; - char id_tag_update[256]; + char id_tag_update[MAX_LIBARRAY]; /* Evaluation context used by viewport */ struct EvaluationContext *eval_ctx; diff --git a/source/blender/blenkernel/BKE_sequencer.h b/source/blender/blenkernel/BKE_sequencer.h index 30bb6954019..811e9136fc9 100644 --- a/source/blender/blenkernel/BKE_sequencer.h +++ b/source/blender/blenkernel/BKE_sequencer.h @@ -477,4 +477,6 @@ struct ImBuf *BKE_sequencer_render_mask_input( int cfra, int fra_offset, bool make_float); void BKE_sequencer_color_balance_apply(struct StripColorBalance *cb, struct ImBuf *ibuf, float mul, bool make_float, struct ImBuf *mask_input); +void BKE_sequencer_all_free_anim_ibufs(int cfra); + #endif /* __BKE_SEQUENCER_H__ */ diff --git a/source/blender/blenkernel/intern/depsgraph.c b/source/blender/blenkernel/intern/depsgraph.c index 85ecf495805..b6bfe336183 100644 --- a/source/blender/blenkernel/intern/depsgraph.c +++ b/source/blender/blenkernel/intern/depsgraph.c @@ -69,6 +69,7 @@ #include "BKE_effect.h" #include "BKE_fcurve.h" #include "BKE_global.h" +#include "BKE_idcode.h" #include "BKE_image.h" #include "BKE_key.h" #include "BKE_library.h" @@ -2602,9 +2603,7 @@ void DAG_ids_flush_tagged(Main *bmain) ListBase *lb = lbarray[a]; ID *id = lb->first; - /* we tag based on first ID type character to avoid - * looping over all ID's in case there are no tags */ - if (id && bmain->id_tag_update[id->name[0]]) { + if (id && bmain->id_tag_update[BKE_idcode_to_index(GS(id->name))]) { for (; id; id = id->next) { if (id->tag & (LIB_TAG_ID_RECALC | LIB_TAG_ID_RECALC_DATA)) { @@ -2644,9 +2643,7 @@ void DAG_ids_check_recalc(Main *bmain, Scene *scene, bool time) ListBase *lb = lbarray[a]; ID *id = lb->first; - /* we tag based on first ID type character to avoid - * looping over all ID's in case there are no tags */ - if (id && bmain->id_tag_update[id->name[0]]) { + if (id && bmain->id_tag_update[BKE_idcode_to_index(GS(id->name))]) { updated = true; break; } @@ -2727,9 +2724,7 @@ void DAG_ids_clear_recalc(Main *bmain) ListBase *lb = lbarray[a]; ID *id = lb->first; - /* we tag based on first ID type character to avoid - * looping over all ID's in case there are no tags */ - if (id && bmain->id_tag_update[id->name[0]]) { + if (id && bmain->id_tag_update[BKE_idcode_to_index(GS(id->name))]) { for (; id; id = id->next) { if (id->tag & (LIB_TAG_ID_RECALC | LIB_TAG_ID_RECALC_DATA)) id->tag &= ~(LIB_TAG_ID_RECALC | LIB_TAG_ID_RECALC_DATA); @@ -2836,12 +2831,12 @@ void DAG_id_type_tag(Main *bmain, short idtype) DAG_id_type_tag(bmain, ID_SCE); } - bmain->id_tag_update[((char *)&idtype)[0]] = 1; + bmain->id_tag_update[BKE_idcode_to_index(idtype)] = 1; } int DAG_id_type_tagged(Main *bmain, short idtype) { - return bmain->id_tag_update[((char *)&idtype)[0]]; + return bmain->id_tag_update[BKE_idcode_to_index(idtype)]; } #if 0 // UNUSED diff --git a/source/blender/blenkernel/intern/idcode.c b/source/blender/blenkernel/intern/idcode.c index ddc4fb37848..c76d072cb64 100644 --- a/source/blender/blenkernel/intern/idcode.c +++ b/source/blender/blenkernel/intern/idcode.c @@ -259,6 +259,55 @@ short BKE_idcode_from_idfilter(const int idfilter) } /** + * Convert an idcode into an index (e.g. ID_OB -> INDEX_ID_OB). + */ +int BKE_idcode_to_index(const short idcode) +{ +#define CASE_IDINDEX(_id) case ID_##_id: return INDEX_ID_##_id + + switch ((ID_Type)idcode) { + CASE_IDINDEX(AC); + CASE_IDINDEX(AR); + CASE_IDINDEX(BR); + CASE_IDINDEX(CA); + CASE_IDINDEX(CF); + CASE_IDINDEX(CU); + CASE_IDINDEX(GD); + CASE_IDINDEX(GR); + CASE_IDINDEX(IM); + CASE_IDINDEX(KE); + CASE_IDINDEX(IP); + CASE_IDINDEX(LA); + CASE_IDINDEX(LI); + CASE_IDINDEX(LS); + CASE_IDINDEX(LT); + CASE_IDINDEX(MA); + CASE_IDINDEX(MB); + CASE_IDINDEX(MC); + CASE_IDINDEX(ME); + CASE_IDINDEX(MSK); + CASE_IDINDEX(NT); + CASE_IDINDEX(OB); + CASE_IDINDEX(PAL); + CASE_IDINDEX(PC); + CASE_IDINDEX(SCE); + CASE_IDINDEX(SCR); + CASE_IDINDEX(SPK); + CASE_IDINDEX(SO); + CASE_IDINDEX(TE); + CASE_IDINDEX(TXT); + CASE_IDINDEX(VF); + CASE_IDINDEX(WM); + CASE_IDINDEX(WO); + } + + BLI_assert(0); + return -1; + +#undef CASE_IDINDEX +} + +/** * Convert an idcode into a name (plural). * * \param idcode: The code to convert. diff --git a/source/blender/blenkernel/intern/library.c b/source/blender/blenkernel/intern/library.c index 933622eed83..ca98f5d04b7 100644 --- a/source/blender/blenkernel/intern/library.c +++ b/source/blender/blenkernel/intern/library.c @@ -744,59 +744,55 @@ void BKE_main_lib_objects_recalc_all(Main *bmain) * \note MAX_LIBARRAY define should match this code */ int set_listbasepointers(Main *main, ListBase **lb) { - int a = 0; - /* BACKWARDS! also watch order of free-ing! (mesh<->mat), first items freed last. * This is important because freeing data decreases usercounts of other datablocks, * if this data is its self freed it can crash. */ - lb[a++] = &(main->library); /* Libraries may be accessed from pretty much any other ID... */ - lb[a++] = &(main->ipo); - lb[a++] = &(main->action); /* moved here to avoid problems when freeing with animato (aligorith) */ - lb[a++] = &(main->key); - lb[a++] = &(main->gpencil); /* referenced by nodes, objects, view, scene etc, before to free after. */ - lb[a++] = &(main->nodetree); - lb[a++] = &(main->image); - lb[a++] = &(main->tex); - lb[a++] = &(main->mat); - lb[a++] = &(main->vfont); + lb[INDEX_ID_LI] = &(main->library); /* Libraries may be accessed from pretty much any other ID... */ + lb[INDEX_ID_IP] = &(main->ipo); + lb[INDEX_ID_AC] = &(main->action); /* moved here to avoid problems when freeing with animato (aligorith) */ + lb[INDEX_ID_KE] = &(main->key); + lb[INDEX_ID_GD] = &(main->gpencil); /* referenced by nodes, objects, view, scene etc, before to free after. */ + lb[INDEX_ID_NT] = &(main->nodetree); + lb[INDEX_ID_IM] = &(main->image); + lb[INDEX_ID_TE] = &(main->tex); + lb[INDEX_ID_MA] = &(main->mat); + lb[INDEX_ID_VF] = &(main->vfont); /* Important!: When adding a new object type, * the specific data should be inserted here */ - lb[a++] = &(main->armature); - - lb[a++] = &(main->cachefiles); - lb[a++] = &(main->mesh); - lb[a++] = &(main->curve); - lb[a++] = &(main->mball); - - lb[a++] = &(main->latt); - lb[a++] = &(main->lamp); - lb[a++] = &(main->camera); - - lb[a++] = &(main->text); - lb[a++] = &(main->sound); - lb[a++] = &(main->group); - lb[a++] = &(main->palettes); - lb[a++] = &(main->paintcurves); - lb[a++] = &(main->brush); - lb[a++] = &(main->speaker); - - lb[a++] = &(main->world); - lb[a++] = &(main->movieclip); - lb[a++] = &(main->screen); - lb[a++] = &(main->object); - lb[a++] = &(main->linestyle); /* referenced by scenes */ - lb[a++] = &(main->scene); - lb[a++] = &(main->wm); - lb[a++] = &(main->mask); + lb[INDEX_ID_AR] = &(main->armature); + + lb[INDEX_ID_CF] = &(main->cachefiles); + lb[INDEX_ID_ME] = &(main->mesh); + lb[INDEX_ID_CU] = &(main->curve); + lb[INDEX_ID_MB] = &(main->mball); + + lb[INDEX_ID_LT] = &(main->latt); + lb[INDEX_ID_LA] = &(main->lamp); + lb[INDEX_ID_CA] = &(main->camera); + + lb[INDEX_ID_TXT] = &(main->text); + lb[INDEX_ID_SO] = &(main->sound); + lb[INDEX_ID_GR] = &(main->group); + lb[INDEX_ID_PAL] = &(main->palettes); + lb[INDEX_ID_PC] = &(main->paintcurves); + lb[INDEX_ID_BR] = &(main->brush); + lb[INDEX_ID_SPK] = &(main->speaker); + + lb[INDEX_ID_WO] = &(main->world); + lb[INDEX_ID_MC] = &(main->movieclip); + lb[INDEX_ID_SCR] = &(main->screen); + lb[INDEX_ID_OB] = &(main->object); + lb[INDEX_ID_LS] = &(main->linestyle); /* referenced by scenes */ + lb[INDEX_ID_SCE] = &(main->scene); + lb[INDEX_ID_WM] = &(main->wm); + lb[INDEX_ID_MSK] = &(main->mask); - lb[a] = NULL; - - BLI_assert(a + 1 == MAX_LIBARRAY); + lb[INDEX_ID_NULL] = NULL; - return a; + return (MAX_LIBARRAY - 1); } /* *********** ALLOC AND FREE ***************** @@ -1645,12 +1641,12 @@ void BKE_library_make_local(Main *bmain, const Library *lib, const bool untagged /* Do not explicitly make local non-linkable IDs (shapekeys, in fact), they are assumed to be handled * by real datablocks responsible of them. */ - const bool do_skip = (id && BKE_idcode_is_linkable(GS(id->name))); + const bool do_skip = (id && !BKE_idcode_is_linkable(GS(id->name))); for (; id; id = id_next) { id->newid = NULL; id_next = id->next; /* id is possibly being inserted again */ - + /* The check on the second line (LIB_TAG_PRE_EXISTING) is done so its * possible to tag data you don't want to be made local, used for * appending data, so any libdata already linked wont become local diff --git a/source/blender/blenkernel/intern/sequencer.c b/source/blender/blenkernel/intern/sequencer.c index 6067a8b2d9b..c240aa27343 100644 --- a/source/blender/blenkernel/intern/sequencer.c +++ b/source/blender/blenkernel/intern/sequencer.c @@ -1656,6 +1656,9 @@ static bool seq_proxy_get_fname(Editing *ed, Sequence *seq, int cfra, int render else if ((proxy->storage & SEQ_STORAGE_PROXY_CUSTOM_DIR) && (proxy->storage & SEQ_STORAGE_PROXY_CUSTOM_FILE)) { BLI_strncpy(dir, seq->strip->proxy->dir, sizeof(dir)); } + else if (proxy->storage & SEQ_STORAGE_PROXY_CUSTOM_FILE) { + BLI_strncpy(dir, seq->strip->proxy->dir, sizeof(dir)); + } else if (sanim && sanim->anim && (proxy->storage & SEQ_STORAGE_PROXY_CUSTOM_DIR)) { char fname[FILE_MAXFILE]; BLI_strncpy(dir, seq->strip->proxy->dir, sizeof(dir)); @@ -1675,13 +1678,21 @@ static bool seq_proxy_get_fname(Editing *ed, Sequence *seq, int cfra, int render if (view_id > 0) BLI_snprintf(suffix, sizeof(suffix), "_%d", view_id); - if (proxy->storage & SEQ_STORAGE_PROXY_CUSTOM_FILE && sanim && sanim->anim && + if (proxy->storage & SEQ_STORAGE_PROXY_CUSTOM_FILE && ed->proxy_storage != SEQ_EDIT_PROXY_DIR_STORAGE) { - BLI_join_dirfile(name, PROXY_MAXFILE, - dir, proxy->file); - BLI_path_abs(name, G.main->name); - BLI_snprintf(name, PROXY_MAXFILE, "%s_%s", name, suffix); + char fname[FILE_MAXFILE]; + BLI_join_dirfile(fname, PROXY_MAXFILE, dir, proxy->file); + BLI_path_abs(fname, G.main->name); + if (suffix[0] != '\0') { + /* TODO(sergey): This will actually append suffix after extension + * which is weird but how was originally coded in multiview branch. + */ + BLI_snprintf(name, PROXY_MAXFILE, "%s_%s", fname, suffix); + } + else { + BLI_strncpy(name, fname, PROXY_MAXFILE); + } return true; } @@ -5595,3 +5606,31 @@ int BKE_sequencer_find_next_prev_edit( return best_frame; } + +static void sequencer_all_free_anim_ibufs(ListBase *seqbase, int cfra) +{ + for (Sequence *seq = seqbase->first; seq != NULL; seq = seq->next) { + if (seq->enddisp < cfra || seq->startdisp > cfra) { + BKE_sequence_free_anim(seq); + } + if (seq->type == SEQ_TYPE_META) { + sequencer_all_free_anim_ibufs(&seq->seqbase, cfra); + } + } +} + +void BKE_sequencer_all_free_anim_ibufs(int cfra) +{ + BKE_sequencer_cache_cleanup(); + for (Scene *scene = G.main->scene.first; + scene != NULL; + scene = scene->id.next) + { + Editing *ed = BKE_sequencer_editing_get(scene, false); + if (ed == NULL) { + /* Ignore scenes without sequencer. */ + continue; + } + sequencer_all_free_anim_ibufs(&ed->seqbase, cfra); + } +} diff --git a/source/blender/depsgraph/intern/depsgraph_query.cc b/source/blender/depsgraph/intern/depsgraph_query.cc index cac4eaae215..7f2f6a65f5e 100644 --- a/source/blender/depsgraph/intern/depsgraph_query.cc +++ b/source/blender/depsgraph/intern/depsgraph_query.cc @@ -33,6 +33,7 @@ #include "MEM_guardedalloc.h" extern "C" { +#include "BKE_idcode.h" #include "BKE_main.h" #include "DEG_depsgraph_query.h" @@ -42,7 +43,7 @@ extern "C" { bool DEG_id_type_tagged(Main *bmain, short idtype) { - return bmain->id_tag_update[((unsigned char *)&idtype)[0]] != 0; + return bmain->id_tag_update[BKE_idcode_to_index(idtype)] != 0; } short DEG_get_eval_flags_for_id(Depsgraph *graph, ID *id) diff --git a/source/blender/depsgraph/intern/depsgraph_tag.cc b/source/blender/depsgraph/intern/depsgraph_tag.cc index 304e3e14bae..75fb55a1b4d 100644 --- a/source/blender/depsgraph/intern/depsgraph_tag.cc +++ b/source/blender/depsgraph/intern/depsgraph_tag.cc @@ -43,6 +43,7 @@ extern "C" { #include "BLI_task.h" +#include "BKE_idcode.h" #include "BKE_library.h" #include "BKE_main.h" #include "BKE_node.h" @@ -225,10 +226,8 @@ void DEG_id_type_tag(Main *bmain, short idtype) DEG_id_type_tag(bmain, ID_WO); DEG_id_type_tag(bmain, ID_SCE); } - /* We tag based on first ID type character to avoid - * looping over all ID's in case there are no tags. - */ - bmain->id_tag_update[((unsigned char *)&idtype)[0]] = 1; + + bmain->id_tag_update[BKE_idcode_to_index(idtype)] = 1; } /* Recursively push updates out to all nodes dependent on this, @@ -339,10 +338,7 @@ void DEG_ids_check_recalc(Main *bmain, Scene *scene, bool time) ListBase *lb = lbarray[a]; ID *id = (ID *)lb->first; - /* We tag based on first ID type character to avoid - * looping over all ID's in case there are no tags. - */ - if (id && bmain->id_tag_update[(unsigned char)id->name[0]]) { + if (id && bmain->id_tag_update[BKE_idcode_to_index(GS(id->name))]) { updated = true; break; } @@ -367,10 +363,7 @@ void DEG_ids_clear_recalc(Main *bmain) ListBase *lb = lbarray[a]; ID *id = (ID *)lb->first; - /* We tag based on first ID type character to avoid - * looping over all ID's in case there are no tags. - */ - if (id && bmain->id_tag_update[(unsigned char)id->name[0]]) { + if (id && bmain->id_tag_update[BKE_idcode_to_index(GS(id->name))]) { for (; id; id = (ID *)id->next) { id->tag &= ~(LIB_TAG_ID_RECALC | LIB_TAG_ID_RECALC_DATA); diff --git a/source/blender/editors/gpencil/gpencil_data.c b/source/blender/editors/gpencil/gpencil_data.c index 6f2ebbed7d7..876f873e5fa 100644 --- a/source/blender/editors/gpencil/gpencil_data.c +++ b/source/blender/editors/gpencil/gpencil_data.c @@ -747,7 +747,7 @@ static int gp_stroke_arrange_exec(bContext *C, wmOperator *op) bGPDframe *gpf = gpl->actframe; /* temp listbase to store selected strokes */ ListBase selected = {NULL}; - const int direction = RNA_enum_get(op->ptr, "type"); + const int direction = RNA_enum_get(op->ptr, "direction"); /* verify if any selected stroke is in the extreme of the stack and select to move */ for (gps = gpf->strokes.first; gps; gps = gps->next) { diff --git a/source/blender/editors/gpencil/gpencil_paint.c b/source/blender/editors/gpencil/gpencil_paint.c index eefdee70304..e7e39a85792 100644 --- a/source/blender/editors/gpencil/gpencil_paint.c +++ b/source/blender/editors/gpencil/gpencil_paint.c @@ -692,6 +692,7 @@ static void gp_stroke_newfrombuffer(tGPsdata *p) bGPDspoint *pt; tGPspoint *ptc; bGPDbrush *brush = p->brush; + ToolSettings *ts = p->scene->toolsettings; int i, totelem; /* since strokes are so fine, when using their depth we need a margin otherwise they might get missed */ @@ -925,8 +926,16 @@ static void gp_stroke_newfrombuffer(tGPsdata *p) gps->palcolor = palcolor; strcpy(gps->colorname, palcolor->info); - /* add stroke to frame */ - BLI_addtail(&p->gpf->strokes, gps); + /* add stroke to frame, usually on tail of the listbase, but if on back is enabled the stroke is added on listbase head + * because the drawing order is inverse and the head stroke is the first to draw. This is very useful for artist + * when drawing the background + */ + if ((ts->gpencil_flags & GP_TOOL_FLAG_PAINT_ONBACK) && (p->paintmode != GP_PAINTMODE_DRAW_POLY)) { + BLI_addhead(&p->gpf->strokes, gps); + } + else { + BLI_addtail(&p->gpf->strokes, gps); + } gp_stroke_added_enable(p); } @@ -2278,6 +2287,28 @@ static void gpencil_stroke_end(wmOperator *op) p->gpf = NULL; } +/* Move last stroke in the listbase to the head to be drawn below all previous strokes in the layer */ +static void gpencil_move_last_stroke_to_back(bContext *C) +{ + /* move last stroke (the polygon) to head of the listbase stroke to draw on back of all previous strokes */ + bGPdata *gpd = ED_gpencil_data_get_active(C); + bGPDlayer *gpl = BKE_gpencil_layer_getactive(gpd); + + /* sanity checks */ + if (ELEM(NULL, gpd, gpl, gpl->actframe)) { + return; + } + + bGPDframe *gpf = gpl->actframe; + bGPDstroke *gps = gpf->strokes.last; + if (ELEM(NULL, gps)) { + return; + } + + BLI_remlink(&gpf->strokes, gps); + BLI_insertlinkbefore(&gpf->strokes, gpf->strokes.first, gps); +} + /* events handling during interactive drawing part of operator */ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) { @@ -2334,6 +2365,12 @@ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) if (ELEM(event->type, RETKEY, PADENTER, ESCKEY, SPACEKEY, EKEY)) { /* exit() ends the current stroke before cleaning up */ /* printf("\t\tGP - end of paint op + end of stroke\n"); */ + /* if drawing polygon and enable on back, must move stroke */ + if ((p->scene->toolsettings->gpencil_flags & GP_TOOL_FLAG_PAINT_ONBACK) && (p->paintmode == GP_PAINTMODE_DRAW_POLY)) { + if (p->flags & GP_PAINTFLAG_STROKEADDED) { + gpencil_move_last_stroke_to_back(C); + } + } p->status = GP_STATUS_DONE; estate = OPERATOR_FINISHED; } @@ -2390,9 +2427,12 @@ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) } else { /* printf("\t\tGP - end of stroke + op\n"); */ - /* disable paint session */ - p->scene->toolsettings->gpencil_flags &= ~GP_TOOL_FLAG_PAINTSESSIONS_ON; - + /* if drawing polygon and enable on back, must move stroke */ + if ((p->scene->toolsettings->gpencil_flags & GP_TOOL_FLAG_PAINT_ONBACK) && (p->paintmode == GP_PAINTMODE_DRAW_POLY)) { + if (p->flags & GP_PAINTFLAG_STROKEADDED) { + gpencil_move_last_stroke_to_back(C); + } + } p->status = GP_STATUS_DONE; estate = OPERATOR_FINISHED; } @@ -2422,9 +2462,6 @@ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) in_bounds = true; } else { - /* disable paint session */ - p->scene->toolsettings->gpencil_flags &= ~GP_TOOL_FLAG_PAINTSESSIONS_ON; - /* Out of bounds, or invalid in some other way */ p->status = GP_STATUS_ERROR; estate = OPERATOR_CANCELLED; @@ -2441,9 +2478,6 @@ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) in_bounds = BLI_rcti_isect_pt_v(®ion_rect, event->mval); } else { - /* disable paint session */ - p->scene->toolsettings->gpencil_flags &= ~GP_TOOL_FLAG_PAINTSESSIONS_ON; - /* No region */ p->status = GP_STATUS_ERROR; estate = OPERATOR_CANCELLED; @@ -2471,9 +2505,6 @@ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) p = gpencil_stroke_begin(C, op); if (p->status == GP_STATUS_ERROR) { - /* disable paint session */ - p->scene->toolsettings->gpencil_flags &= ~GP_TOOL_FLAG_PAINTSESSIONS_ON; - estate = OPERATOR_CANCELLED; } } @@ -2482,9 +2513,12 @@ static int gpencil_draw_modal(bContext *C, wmOperator *op, const wmEvent *event) * NOTE: Don't eter this case if an error occurred while finding the * region (as above) */ - /* disable paint session */ - p->scene->toolsettings->gpencil_flags &= ~GP_TOOL_FLAG_PAINTSESSIONS_ON; - + /* if drawing polygon and enable on back, must move stroke */ + if ((p->scene->toolsettings->gpencil_flags & GP_TOOL_FLAG_PAINT_ONBACK) && (p->paintmode == GP_PAINTMODE_DRAW_POLY)) { + if (p->flags & GP_PAINTFLAG_STROKEADDED) { + gpencil_move_last_stroke_to_back(C); + } + } p->status = GP_STATUS_DONE; estate = OPERATOR_FINISHED; } diff --git a/source/blender/editors/interface/interface_templates.c b/source/blender/editors/interface/interface_templates.c index b3f2192c26c..bb9da1a141b 100644 --- a/source/blender/editors/interface/interface_templates.c +++ b/source/blender/editors/interface/interface_templates.c @@ -857,8 +857,8 @@ static uiLayout *draw_modifier( /* mode enabling buttons */ UI_block_align_begin(block); - /* Softbody not allowed in this situation, enforce! */ - if (((md->type != eModifierType_Softbody && md->type != eModifierType_Collision) || !(ob->pd && ob->pd->deflect)) && + /* Collision and Surface are always enabled, hide buttons! */ + if (((md->type != eModifierType_Collision) || !(ob->pd && ob->pd->deflect)) && (md->type != eModifierType_Surface) ) { uiItemR(row, &ptr, "show_render", 0, "", ICON_NONE); diff --git a/source/blender/editors/object/object_relations.c b/source/blender/editors/object/object_relations.c index 6a9f5a9bfaf..fb067ccca81 100644 --- a/source/blender/editors/object/object_relations.c +++ b/source/blender/editors/object/object_relations.c @@ -1381,7 +1381,7 @@ static int move_to_layer_exec(bContext *C, wmOperator *op) /* upper byte is used for local view */ local = base->lay & 0xFF000000; base->lay = lay + local; - base->object->lay = lay; + base->object->lay = base->lay; /* if (base->object->type == OB_LAMP) is_lamp = true; */ } CTX_DATA_END; diff --git a/source/blender/editors/space_graph/space_graph.c b/source/blender/editors/space_graph/space_graph.c index 2582ba4be8d..b35d1b2c777 100644 --- a/source/blender/editors/space_graph/space_graph.c +++ b/source/blender/editors/space_graph/space_graph.c @@ -687,7 +687,7 @@ static void graph_id_remap(ScrArea *UNUSED(sa), SpaceLink *slink, ID *old_id, ID return; } - if ((ID *)sgraph->ads->filter_grp == old_id) { + if (sgraph->ads && (ID *)sgraph->ads->filter_grp == old_id) { sgraph->ads->filter_grp = (Group *)new_id; } } diff --git a/source/blender/editors/space_view3d/drawvolume.c b/source/blender/editors/space_view3d/drawvolume.c index a667b46d676..29c06388cf6 100644 --- a/source/blender/editors/space_view3d/drawvolume.c +++ b/source/blender/editors/space_view3d/drawvolume.c @@ -378,11 +378,13 @@ void draw_smoke_volume(SmokeDomainSettings *sds, Object *ob, /* setup buffer and draw */ - int gl_depth = 0, gl_blend = 0; + int gl_depth = 0, gl_blend = 0, gl_depth_write = 0; glGetBooleanv(GL_BLEND, (GLboolean *)&gl_blend); glGetBooleanv(GL_DEPTH_TEST, (GLboolean *)&gl_depth); + glGetBooleanv(GL_DEPTH_WRITEMASK, (GLboolean *)&gl_depth_write); glEnable(GL_DEPTH_TEST); + glDepthMask(GL_FALSE); glEnable(GL_BLEND); glBlendFunc(GL_ONE, GL_ONE_MINUS_SRC_ALPHA); @@ -421,6 +423,8 @@ void draw_smoke_volume(SmokeDomainSettings *sds, Object *ob, GPU_shader_unbind(); + glDepthMask(gl_depth_write); + if (!gl_blend) { glDisable(GL_BLEND); } diff --git a/source/blender/gpu/intern/gpu_texture.c b/source/blender/gpu/intern/gpu_texture.c index 827c52c9a5f..54f0003c086 100644 --- a/source/blender/gpu/intern/gpu_texture.c +++ b/source/blender/gpu/intern/gpu_texture.c @@ -55,7 +55,8 @@ struct GPUTexture { int number; /* number for multitexture binding */ int refcount; /* reference count */ GLenum target; /* GL_TEXTURE_* */ - GLenum target_base; /* same as target, (but no multisample) */ + GLenum target_base; /* same as target, (but no multisample) + * use it for unbinding */ GLuint bindcode; /* opengl identifier for texture */ int fromblender; /* we got the texture from Blender */ @@ -374,6 +375,9 @@ GPUTexture *GPU_texture_from_blender(Image *ima, ImageUser *iuser, int textarget GLint bindcode = GPU_verify_image(ima, iuser, textarget, 0, 0, mipmap, is_data); GPU_update_image_time(ima, time); + /* see GPUInput::textarget: it can take two values - GL_TEXTURE_2D and GL_TEXTURE_CUBE_MAP + * these values are correct for glDisable, so textarget can be safely used in + * GPU_texture_bind/GPU_texture_unbind through tex->target_base */ if (textarget == GL_TEXTURE_2D) gputt = TEXTARGET_TEXTURE_2D; else @@ -390,7 +394,7 @@ GPUTexture *GPU_texture_from_blender(Image *ima, ImageUser *iuser, int textarget tex->number = -1; tex->refcount = 1; tex->target = textarget; - tex->target_base = GL_TEXTURE_2D; + tex->target_base = textarget; tex->fromblender = 1; ima->gputexture[gputt] = tex; @@ -626,11 +630,11 @@ void GPU_texture_bind(GPUTexture *tex, int number) GLenum arbnumber = (GLenum)((GLuint)GL_TEXTURE0 + number); if (number != 0) glActiveTexture(arbnumber); if (tex->bindcode != 0) { - glBindTexture(tex->target, tex->bindcode); + glBindTexture(tex->target_base, tex->bindcode); } else - GPU_invalid_tex_bind(tex->target); - glEnable(tex->target); + GPU_invalid_tex_bind(tex->target_base); + glEnable(tex->target_base); if (number != 0) glActiveTexture(GL_TEXTURE0); tex->number = number; @@ -652,8 +656,6 @@ void GPU_texture_unbind(GPUTexture *tex) GLenum arbnumber = (GLenum)((GLuint)GL_TEXTURE0 + tex->number); if (tex->number != 0) glActiveTexture(arbnumber); - glBindTexture(tex->target, 0); - glDisable(tex->target); glBindTexture(tex->target_base, 0); glDisable(tex->target_base); if (tex->number != 0) glActiveTexture(GL_TEXTURE0); diff --git a/source/blender/makesdna/DNA_ID.h b/source/blender/makesdna/DNA_ID.h index 725eefe5155..45eb31235f5 100644 --- a/source/blender/makesdna/DNA_ID.h +++ b/source/blender/makesdna/DNA_ID.h @@ -379,6 +379,45 @@ enum { FILTER_ID_CF = (1 << 27), }; +/* IMPORTANT: this enum matches the order currently use in set_lisbasepointers, + * keep them in sync! */ +enum { + INDEX_ID_LI = 0, + INDEX_ID_IP, + INDEX_ID_AC, + INDEX_ID_KE, + INDEX_ID_GD, + INDEX_ID_NT, + INDEX_ID_IM, + INDEX_ID_TE, + INDEX_ID_MA, + INDEX_ID_VF, + INDEX_ID_AR, + INDEX_ID_CF, + INDEX_ID_ME, + INDEX_ID_CU, + INDEX_ID_MB, + INDEX_ID_LT, + INDEX_ID_LA, + INDEX_ID_CA, + INDEX_ID_TXT, + INDEX_ID_SO, + INDEX_ID_GR, + INDEX_ID_PAL, + INDEX_ID_PC, + INDEX_ID_BR, + INDEX_ID_SPK, + INDEX_ID_WO, + INDEX_ID_MC, + INDEX_ID_SCR, + INDEX_ID_OB, + INDEX_ID_LS, + INDEX_ID_SCE, + INDEX_ID_WM, + INDEX_ID_MSK, + INDEX_ID_NULL, +}; + #ifdef __cplusplus } #endif diff --git a/source/blender/makesdna/DNA_scene_types.h b/source/blender/makesdna/DNA_scene_types.h index 66b79282533..d4edc8a1c20 100644 --- a/source/blender/makesdna/DNA_scene_types.h +++ b/source/blender/makesdna/DNA_scene_types.h @@ -2052,6 +2052,8 @@ typedef enum eGPencil_Flags { GP_TOOL_FLAG_PAINTSESSIONS_ON = (1 << 0), /* When creating new frames, the last frame gets used as the basis for the new one */ GP_TOOL_FLAG_RETAIN_LAST = (1 << 1), + /* Add the strokes below all strokes in the layer */ + GP_TOOL_FLAG_PAINT_ONBACK = (1 << 2) } eGPencil_Flags; /* toolsettings->gpencil_src */ diff --git a/source/blender/makesrna/intern/rna_gpencil.c b/source/blender/makesrna/intern/rna_gpencil.c index 80f4d5d05b6..7424c190501 100644 --- a/source/blender/makesrna/intern/rna_gpencil.c +++ b/source/blender/makesrna/intern/rna_gpencil.c @@ -1381,8 +1381,9 @@ static void rna_def_gpencil_palettecolor(BlenderRNA *brna) RNA_def_property_update(prop, NC_GPENCIL | ND_DATA, "rna_GPencil_update"); /* Name */ - prop = RNA_def_property(srna, "info", PROP_STRING, PROP_NONE); - RNA_def_property_ui_text(prop, "Info", "Color name"); + prop = RNA_def_property(srna, "name", PROP_STRING, PROP_NONE); + RNA_def_property_string_sdna(prop, NULL, "info"); + RNA_def_property_ui_text(prop, "Name", "Color name"); RNA_def_property_string_funcs(prop, NULL, NULL, "rna_GPencilPaletteColor_info_set"); RNA_def_struct_name_property(srna, prop); RNA_def_property_update(prop, NC_GPENCIL | ND_DATA, "rna_GPencil_update"); diff --git a/source/blender/makesrna/intern/rna_scene.c b/source/blender/makesrna/intern/rna_scene.c index f39e7dc10b7..06b5bd87dd4 100644 --- a/source/blender/makesrna/intern/rna_scene.c +++ b/source/blender/makesrna/intern/rna_scene.c @@ -2593,6 +2593,12 @@ static void rna_def_tool_settings(BlenderRNA *brna) "are included as the basis for the new one"); RNA_def_property_update(prop, NC_SCENE | ND_TOOLSETTINGS, NULL); + prop = RNA_def_property(srna, "use_gpencil_draw_onback", PROP_BOOLEAN, PROP_NONE); + RNA_def_property_boolean_sdna(prop, NULL, "gpencil_flags", GP_TOOL_FLAG_PAINT_ONBACK); + RNA_def_property_ui_text(prop, "Draw Strokes on Back", + "When draw new strokes, the new stroke is drawn below of all strokes in the layer"); + RNA_def_property_update(prop, NC_SCENE | ND_TOOLSETTINGS, NULL); + prop = RNA_def_property(srna, "grease_pencil_source", PROP_ENUM, PROP_NONE); RNA_def_property_enum_bitflag_sdna(prop, NULL, "gpencil_src"); RNA_def_property_enum_items(prop, gpencil_source_3d_items); diff --git a/source/blender/render/intern/raytrace/rayobject_rtbuild.cpp b/source/blender/render/intern/raytrace/rayobject_rtbuild.cpp index 02a49fc3c8f..724a809077e 100644 --- a/source/blender/render/intern/raytrace/rayobject_rtbuild.cpp +++ b/source/blender/render/intern/raytrace/rayobject_rtbuild.cpp @@ -42,6 +42,10 @@ #include "BLI_math.h" #include "BLI_utildefines.h" +#if __cplusplus >= 201103L +using std::isfinite; +#endif + static bool selected_node(RTBuilder::Object *node) { return node->selected; diff --git a/source/blender/render/intern/source/pipeline.c b/source/blender/render/intern/source/pipeline.c index fc99bedfa0f..363267bb363 100644 --- a/source/blender/render/intern/source/pipeline.c +++ b/source/blender/render/intern/source/pipeline.c @@ -2784,6 +2784,7 @@ static void do_render_all_options(Render *re) /* ensure no images are in memory from previous animated sequences */ BKE_image_all_free_anim_ibufs(re->r.cfra); + BKE_sequencer_all_free_anim_ibufs(re->r.cfra); if (RE_engine_render(re, 1)) { /* in this case external render overrides all */ diff --git a/source/gameengine/GamePlayer/ghost/GPG_ghost.cpp b/source/gameengine/GamePlayer/ghost/GPG_ghost.cpp index d033afacc08..edbbf93bf7a 100644 --- a/source/gameengine/GamePlayer/ghost/GPG_ghost.cpp +++ b/source/gameengine/GamePlayer/ghost/GPG_ghost.cpp @@ -61,6 +61,7 @@ extern "C" #include "DNA_scene_types.h" #include "DNA_userdef_types.h" +#include "DNA_genfile.h" #include "BLO_readfile.h" #include "BLO_runtime.h" @@ -492,6 +493,8 @@ int main( // freeing up GPU_Textures works correctly. BLI_threadapi_init(); + DNA_sdna_current_init(); + RNA_init(); init_nodesystem(); |