blender
/
blender
136
404
Fork 586
Code Issues 6.2k Pull Requests 759 Projects 19 Wiki Activity

Merge branch 'master' into blender2.8

This commit is contained in:
Campbell Barton 2017-05-20 14:19:05 +10:00
parent 996bf65730 81e584ed17
commit 65aab6cdae
82 changed files with 513 additions and 383 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

7
CMakeLists.txt
View File

@ -1347,9 +1347,16 @@ if(CMAKE_COMPILER_IS_GNUCC)
ADD_CHECK_CXX_COMPILER_FLAG(CXX_WARNINGS CXX_WARN_MISSING_DECLARATIONS -Wmissing-declarations)
endif()
# Use 'ATTR_FALLTHROUGH' macro to suppress.
if(CMAKE_COMPILER_IS_GNUCC AND (NOT "${CMAKE_C_COMPILER_VERSION}" VERSION_LESS "7.0"))
ADD_CHECK_C_COMPILER_FLAG(C_WARNINGS C_WARN_IMPLICIT_FALLTHROUGH -Wimplicit-fallthrough=5)
ADD_CHECK_CXX_COMPILER_FLAG(CXX_WARNINGS CXX_WARN_IMPLICIT_FALLTHROUGH -Wimplicit-fallthrough=5)
endif()
# flags to undo strict flags
ADD_CHECK_C_COMPILER_FLAG(CC_REMOVE_STRICT_FLAGS C_WARN_NO_DEPRECATED_DECLARATIONS -Wno-deprecated-declarations)
ADD_CHECK_C_COMPILER_FLAG(CC_REMOVE_STRICT_FLAGS C_WARN_NO_UNUSED_PARAMETER -Wno-unused-parameter)
ADD_CHECK_C_COMPILER_FLAG(CC_REMOVE_STRICT_FLAGS C_WARN_NO_IMPLICIT_FALLTHROUGH -Wno-implicit-fallthrough)
if(NOT APPLE)
ADD_CHECK_C_COMPILER_FLAG(CC_REMOVE_STRICT_FLAGS C_WARN_NO_ERROR_UNUSED_BUT_SET_VARIABLE -Wno-error=unused-but-set-variable)

26
intern/cycles/kernel/filter/filter_features.h
View File

@ -28,7 +28,11 @@
pixel_buffer += buffer_w - (high.x - low.x); \
}
ccl_device_inline void filter_get_features(int2 pixel, ccl_global float ccl_restrict_ptr buffer, float *features, float ccl_restrict_ptr mean, int pass_stride)
ccl_device_inline void filter_get_features(int2 pixel,
const ccl_global float *ccl_restrict buffer,
float *features,
const float *ccl_restrict mean,
int pass_stride)
{
features[0] = pixel.x;
features[1] = pixel.y;
@ -46,7 +50,11 @@ ccl_device_inline void filter_get_features(int2 pixel, ccl_global float ccl_rest
}
}
ccl_device_inline void filter_get_feature_scales(int2 pixel, ccl_global float ccl_restrict_ptr buffer, float *scales, float ccl_restrict_ptr mean, int pass_stride)
ccl_device_inline void filter_get_feature_scales(int2 pixel,
const ccl_global float *ccl_restrict buffer,
float *scales,
const float *ccl_restrict mean,
int pass_stride)
{
scales[0] = fabsf(pixel.x - mean[0]);
scales[1] = fabsf(pixel.y - mean[1]);
@ -70,19 +78,21 @@ ccl_device_inline void filter_calculate_scale(float *scale)
scale[3] = scale[4] = scale[5] = 1.0f/max(sqrtf(scale[3]), 0.01f);
}
ccl_device_inline float3 filter_get_pixel_color(ccl_global float ccl_restrict_ptr buffer, int pass_stride)
ccl_device_inline float3 filter_get_pixel_color(const ccl_global float *ccl_restrict buffer,
int pass_stride)
{
return make_float3(ccl_get_feature(buffer, 0), ccl_get_feature(buffer, 1), ccl_get_feature(buffer, 2));
}
ccl_device_inline float filter_get_pixel_variance(ccl_global float ccl_restrict_ptr buffer, int pass_stride)
ccl_device_inline float filter_get_pixel_variance(const ccl_global float *ccl_restrict buffer,
int pass_stride)
{
return average(make_float3(ccl_get_feature(buffer, 0), ccl_get_feature(buffer, 1), ccl_get_feature(buffer, 2)));
}
ccl_device_inline void design_row_add(float *design_row,
int rank,
ccl_global float ccl_restrict_ptr transform,
const ccl_global float *ccl_restrict transform,
int stride,
int row,
float feature)
@ -94,13 +104,13 @@ ccl_device_inline void design_row_add(float *design_row,
/* Fill the design row. */
ccl_device_inline void filter_get_design_row_transform(int2 p_pixel,
ccl_global float ccl_restrict_ptr p_buffer,
const ccl_global float *ccl_restrict p_buffer,
int2 q_pixel,
ccl_global float ccl_restrict_ptr q_buffer,
const ccl_global float *ccl_restrict q_buffer,
int pass_stride,
int rank,
float *design_row,
ccl_global float ccl_restrict_ptr transform,
const ccl_global float *ccl_restrict transform,
int stride)
{
design_row[0] = 1.0f;

14
intern/cycles/kernel/filter/filter_features_sse.h
View File

@ -33,7 +33,12 @@ CCL_NAMESPACE_BEGIN
pixel_buffer += buffer_w - (pixel.x - low.x); \
}
ccl_device_inline void filter_get_features_sse(__m128 x, __m128 y, __m128 active_pixels, float ccl_restrict_ptr buffer, __m128 *features, __m128 ccl_restrict_ptr mean, int pass_stride)
ccl_device_inline void filter_get_features_sse(__m128 x, __m128 y,
__m128 active_pixels,
const float *ccl_restrict buffer,
__m128 *features,
const __m128 *ccl_restrict mean,
int pass_stride)
{
features[0] = x;
features[1] = y;
@ -53,7 +58,12 @@ ccl_device_inline void filter_get_features_sse(__m128 x, __m128 y, __m128 active
features[i] = _mm_mask_ps(features[i], active_pixels);
}
ccl_device_inline void filter_get_feature_scales_sse(__m128 x, __m128 y, __m128 active_pixels, float ccl_restrict_ptr buffer, __m128 *scales, __m128 ccl_restrict_ptr mean, int pass_stride)
ccl_device_inline void filter_get_feature_scales_sse(__m128 x, __m128 y,
__m128 active_pixels,
const float *ccl_restrict buffer,
__m128 *scales,
const __m128 *ccl_restrict mean,
int pass_stride)
{
scales[0] = _mm_mask_ps(_mm_fabs_ps(_mm_sub_ps(x, mean[0])), active_pixels);
scales[1] = _mm_mask_ps(_mm_fabs_ps(_mm_sub_ps(y, mean[1])), active_pixels);

72
intern/cycles/kernel/filter/filter_nlm_cpu.h
View File

@ -16,27 +16,39 @@
CCL_NAMESPACE_BEGIN
ccl_device_inline void kernel_filter_nlm_calc_difference(int dx, int dy, float ccl_restrict_ptr weightImage, float ccl_restrict_ptr varianceImage, float *differenceImage, int4 rect, int w, int channel_offset, float a, float k_2)
ccl_device_inline void kernel_filter_nlm_calc_difference(int dx, int dy,
const float *ccl_restrict weight_image,
const float *ccl_restrict variance_image,
float *difference_image,
int4 rect,
int w,
int channel_offset,
float a,
float k_2)
{
for(int y = rect.y; y < rect.w; y++) {
for(int x = rect.x; x < rect.z; x++) {
float diff = 0.0f;
int numChannels = channel_offset? 3 : 1;
for(int c = 0; c < numChannels; c++) {
float cdiff = weightImage[c*channel_offset + y*w+x] - weightImage[c*channel_offset + (y+dy)*w+(x+dx)];
float pvar = varianceImage[c*channel_offset + y*w+x];
float qvar = varianceImage[c*channel_offset + (y+dy)*w+(x+dx)];
float cdiff = weight_image[c*channel_offset + y*w+x] - weight_image[c*channel_offset + (y+dy)*w+(x+dx)];
float pvar = variance_image[c*channel_offset + y*w+x];
float qvar = variance_image[c*channel_offset + (y+dy)*w+(x+dx)];
diff += (cdiff*cdiff - a*(pvar + min(pvar, qvar))) / (1e-8f + k_2*(pvar+qvar));
}
if(numChannels > 1) {
diff *= 1.0f/numChannels;
}
differenceImage[y*w+x] = diff;
difference_image[y*w+x] = diff;
}
}
}
ccl_device_inline void kernel_filter_nlm_blur(float ccl_restrict_ptr differenceImage, float *outImage, int4 rect, int w, int f)
ccl_device_inline void kernel_filter_nlm_blur(const float *ccl_restrict difference_image,
float *out_image,
int4 rect,
int w,
int f)
{
#ifdef __KERNEL_SSE3__
int aligned_lowx = (rect.x & ~(3));
@ -46,30 +58,34 @@ ccl_device_inline void kernel_filter_nlm_blur(float ccl_restrict_ptr differenceI
const int low = max(rect.y, y-f);
const int high = min(rect.w, y+f+1);
for(int x = rect.x; x < rect.z; x++) {
outImage[y*w+x] = 0.0f;
out_image[y*w+x] = 0.0f;
}
for(int y1 = low; y1 < high; y1++) {
#ifdef __KERNEL_SSE3__
for(int x = aligned_lowx; x < aligned_highx; x+=4) {
_mm_store_ps(outImage + y*w+x, _mm_add_ps(_mm_load_ps(outImage + y*w+x), _mm_load_ps(differenceImage + y1*w+x)));
_mm_store_ps(out_image + y*w+x, _mm_add_ps(_mm_load_ps(out_image + y*w+x), _mm_load_ps(difference_image + y1*w+x)));
}
#else
for(int x = rect.x; x < rect.z; x++) {
outImage[y*w+x] += differenceImage[y1*w+x];
out_image[y*w+x] += difference_image[y1*w+x];
}
#endif
}
for(int x = rect.x; x < rect.z; x++) {
outImage[y*w+x] *= 1.0f/(high - low);
out_image[y*w+x] *= 1.0f/(high - low);
}
}
}
ccl_device_inline void kernel_filter_nlm_calc_weight(float ccl_restrict_ptr differenceImage, float *outImage, int4 rect, int w, int f)
ccl_device_inline void kernel_filter_nlm_calc_weight(const float *ccl_restrict difference_image,
float *out_image,
int4 rect,
int w,
int f)
{
for(int y = rect.y; y < rect.w; y++) {
for(int x = rect.x; x < rect.z; x++) {
outImage[y*w+x] = 0.0f;
out_image[y*w+x] = 0.0f;
}
}
for(int dx = -f; dx <= f; dx++) {
@ -77,7 +93,7 @@ ccl_device_inline void kernel_filter_nlm_calc_weight(float ccl_restrict_ptr diff
int neg_dx = min(0, dx);
for(int y = rect.y; y < rect.w; y++) {
for(int x = rect.x-neg_dx; x < rect.z-pos_dx; x++) {
outImage[y*w+x] += differenceImage[y*w+dx+x];
out_image[y*w+x] += difference_image[y*w+dx+x];
}
}
}
@ -85,12 +101,19 @@ ccl_device_inline void kernel_filter_nlm_calc_weight(float ccl_restrict_ptr diff
for(int x = rect.x; x < rect.z; x++) {
const int low = max(rect.x, x-f);
const int high = min(rect.z, x+f+1);
outImage[y*w+x] = expf(-max(outImage[y*w+x] * (1.0f/(high - low)), 0.0f));
out_image[y*w+x] = expf(-max(out_image[y*w+x] * (1.0f/(high - low)), 0.0f));
}
}
}
ccl_device_inline void kernel_filter_nlm_update_output(int dx, int dy, float ccl_restrict_ptr differenceImage, float ccl_restrict_ptr image, float *outImage, float *accumImage, int4 rect, int w, int f)
ccl_device_inline void kernel_filter_nlm_update_output(int dx, int dy,
const float *ccl_restrict difference_image,
const float *ccl_restrict image,
float *out_image,
float *accum_image,
int4 rect,
int w,
int f)
{
for(int y = rect.y; y < rect.w; y++) {
for(int x = rect.x; x < rect.z; x++) {
@ -98,18 +121,18 @@ ccl_device_inline void kernel_filter_nlm_update_output(int dx, int dy, float ccl
const int high = min(rect.z, x+f+1);
float sum = 0.0f;
for(int x1 = low; x1 < high; x1++) {
sum += differenceImage[y*w+x1];
sum += difference_image[y*w+x1];
}
float weight = sum * (1.0f/(high - low));
accumImage[y*w+x] += weight;
outImage[y*w+x] += weight*image[(y+dy)*w+(x+dx)];
accum_image[y*w+x] += weight;
out_image[y*w+x] += weight*image[(y+dy)*w+(x+dx)];
}
}
}
ccl_device_inline void kernel_filter_nlm_construct_gramian(int dx, int dy,
float ccl_restrict_ptr differenceImage,
float ccl_restrict_ptr buffer,
const float *ccl_restrict difference_image,
const float *ccl_restrict buffer,
float *color_pass,
float *variance_pass,
float *transform,
@ -130,7 +153,7 @@ ccl_device_inline void kernel_filter_nlm_construct_gramian(int dx, int dy,
const int high = min(rect.z, x+f+1);
float sum = 0.0f;
for(int x1 = low; x1 < high; x1++) {
sum += differenceImage[y*w+x1];
sum += difference_image[y*w+x1];
}
float weight = sum * (1.0f/(high - low));
@ -151,11 +174,14 @@ ccl_device_inline void kernel_filter_nlm_construct_gramian(int dx, int dy,
}
}
ccl_device_inline void kernel_filter_nlm_normalize(float *outImage, float ccl_restrict_ptr accumImage, int4 rect, int w)
ccl_device_inline void kernel_filter_nlm_normalize(float *out_image,
const float *ccl_restrict accum_image,
int4 rect,
int w)
{
for(int y = rect.y; y < rect.w; y++) {
for(int x = rect.x; x < rect.z; x++) {
outImage[y*w+x] /= accumImage[y*w+x];
out_image[y*w+x] /= accum_image[y*w+x];
}
}
}

62
intern/cycles/kernel/filter/filter_nlm_gpu.h
View File

@ -18,9 +18,9 @@ CCL_NAMESPACE_BEGIN
ccl_device_inline void kernel_filter_nlm_calc_difference(int x, int y,
int dx, int dy,
ccl_global float ccl_restrict_ptr weightImage,
ccl_global float ccl_restrict_ptr varianceImage,
ccl_global float *differenceImage,
const ccl_global float *ccl_restrict weight_image,
const ccl_global float *ccl_restrict variance_image,
ccl_global float *difference_image,
int4 rect, int w,
int channel_offset,
float a, float k_2)
@ -28,78 +28,78 @@ ccl_device_inline void kernel_filter_nlm_calc_difference(int x, int y,
float diff = 0.0f;
int numChannels = channel_offset? 3 : 1;
for(int c = 0; c < numChannels; c++) {
float cdiff = weightImage[c*channel_offset + y*w+x] - weightImage[c*channel_offset + (y+dy)*w+(x+dx)];
float pvar = varianceImage[c*channel_offset + y*w+x];
float qvar = varianceImage[c*channel_offset + (y+dy)*w+(x+dx)];
float cdiff = weight_image[c*channel_offset + y*w+x] - weight_image[c*channel_offset + (y+dy)*w+(x+dx)];
float pvar = variance_image[c*channel_offset + y*w+x];
float qvar = variance_image[c*channel_offset + (y+dy)*w+(x+dx)];
diff += (cdiff*cdiff - a*(pvar + min(pvar, qvar))) / (1e-8f + k_2*(pvar+qvar));
}
if(numChannels > 1) {
diff *= 1.0f/numChannels;
}
differenceImage[y*w+x] = diff;
difference_image[y*w+x] = diff;
}
ccl_device_inline void kernel_filter_nlm_blur(int x, int y,
ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float *outImage,
const ccl_global float *ccl_restrict difference_image,
ccl_global float *out_image,
int4 rect, int w, int f)
{
float sum = 0.0f;
const int low = max(rect.y, y-f);
const int high = min(rect.w, y+f+1);
for(int y1 = low; y1 < high; y1++) {
sum += differenceImage[y1*w+x];
sum += difference_image[y1*w+x];
}
sum *= 1.0f/(high-low);
outImage[y*w+x] = sum;
out_image[y*w+x] = sum;
}
ccl_device_inline void kernel_filter_nlm_calc_weight(int x, int y,
ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float *outImage,
const ccl_global float *ccl_restrict difference_image,
ccl_global float *out_image,
int4 rect, int w, int f)
{
float sum = 0.0f;
const int low = max(rect.x, x-f);
const int high = min(rect.z, x+f+1);
for(int x1 = low; x1 < high; x1++) {
sum += differenceImage[y*w+x1];
sum += difference_image[y*w+x1];
}
sum *= 1.0f/(high-low);
outImage[y*w+x] = expf(-max(sum, 0.0f));
out_image[y*w+x] = expf(-max(sum, 0.0f));
}
ccl_device_inline void kernel_filter_nlm_update_output(int x, int y,
int dx, int dy,
ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float ccl_restrict_ptr image,
ccl_global float *outImage,
ccl_global float *accumImage,
const ccl_global float *ccl_restrict difference_image,
const ccl_global float *ccl_restrict image,
ccl_global float *out_image,
ccl_global float *accum_image,
int4 rect, int w, int f)
{
float sum = 0.0f;
const int low = max(rect.x, x-f);
const int high = min(rect.z, x+f+1);
for(int x1 = low; x1 < high; x1++) {
sum += differenceImage[y*w+x1];
sum += difference_image[y*w+x1];
}
sum *= 1.0f/(high-low);
if(outImage) {
accumImage[y*w+x] += sum;
outImage[y*w+x] += sum*image[(y+dy)*w+(x+dx)];
if(out_image) {
accum_image[y*w+x] += sum;
out_image[y*w+x] += sum*image[(y+dy)*w+(x+dx)];
}
else {
accumImage[y*w+x] = sum;
accum_image[y*w+x] = sum;
}
}
ccl_device_inline void kernel_filter_nlm_construct_gramian(int fx, int fy,
int dx, int dy,
ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float ccl_restrict_ptr buffer,
const ccl_global float *ccl_restrict difference_image,
const ccl_global float *ccl_restrict buffer,
ccl_global float *color_pass,
ccl_global float *variance_pass,
ccl_global float ccl_restrict_ptr transform,
const ccl_global float *ccl_restrict transform,
ccl_global int *rank,
ccl_global float *XtWX,
ccl_global float3 *XtWY,
@ -115,7 +115,7 @@ ccl_device_inline void kernel_filter_nlm_construct_gramian(int fx, int fy,
const int high = min(rect.z, x+f+1);
float sum = 0.0f;
for(int x1 = low; x1 < high; x1++) {
sum += differenceImage[y*w+x1];
sum += difference_image[y*w+x1];
}
float weight = sum * (1.0f/(high - low));
@ -137,11 +137,11 @@ ccl_device_inline void kernel_filter_nlm_construct_gramian(int fx, int fy,
}
ccl_device_inline void kernel_filter_nlm_normalize(int x, int y,
ccl_global float *outImage,
ccl_global float ccl_restrict_ptr accumImage,
ccl_global float *out_image,
const ccl_global float *ccl_restrict accum_image,
int4 rect, int w)
{
outImage[y*w+x] /= accumImage[y*w+x];
out_image[y*w+x] /= accum_image[y*w+x];
}
CCL_NAMESPACE_END

2
intern/cycles/kernel/filter/filter_prefilter.h
View File

@ -44,7 +44,7 @@ ccl_device void kernel_filter_divide_shadow(int sample,
int offset = tiles->offsets[tile];
int stride = tiles->strides[tile];
ccl_global float ccl_restrict_ptr center_buffer = (ccl_global float*) tiles->buffers[tile];
const ccl_global float *ccl_restrict center_buffer = (ccl_global float*) tiles->buffers[tile];
center_buffer += (y*stride + x + offset)*buffer_pass_stride;
center_buffer += buffer_denoising_offset + 14;

12
intern/cycles/kernel/filter/filter_reconstruction.h
View File

@ -21,10 +21,10 @@ ccl_device_inline void kernel_filter_construct_gramian(int x, int y,
int dx, int dy,
int w, int h,
int pass_stride,
ccl_global float ccl_restrict_ptr buffer,
const ccl_global float *ccl_restrict buffer,
ccl_global float *color_pass,
ccl_global float *variance_pass,
ccl_global float ccl_restrict_ptr transform,
const ccl_global float *ccl_restrict transform,
ccl_global int *rank,
float weight,
ccl_global float *XtWX,
@ -85,9 +85,17 @@ ccl_device_inline void kernel_filter_finalize(int x, int y, int w, int h,
const int stride = storage_stride;
#endif
/* The weighted average of pixel colors (essentially, the NLM-filtered image).
* In case the solution of the linear model fails due to numerical issues,
* fall back to this value. */
float3 mean_color = XtWY[0]/XtWX[0];
math_trimatrix_vec3_solve(XtWX, XtWY, (*rank)+1, stride);
float3 final_color = XtWY[0];
if(!isfinite3_safe(final_color)) {
final_color = mean_color;
}
ccl_global float *combined_buffer = buffer + (y*buffer_params.y + x + buffer_params.x)*buffer_params.z;
final_color *= sample;

21
intern/cycles/kernel/filter/filter_transform.h
View File

@ -16,7 +16,7 @@
CCL_NAMESPACE_BEGIN
ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
ccl_device void kernel_filter_construct_transform(const float *ccl_restrict buffer,
int x, int y, int4 rect,
int pass_stride,
float *transform, int *rank,
@ -29,20 +29,15 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
/* Temporary storage, used in different steps of the algorithm. */
float tempmatrix[DENOISE_FEATURES*DENOISE_FEATURES];
float tempvector[2*DENOISE_FEATURES];
float ccl_restrict_ptr pixel_buffer;
const float *ccl_restrict pixel_buffer;
int2 pixel;
/* === Calculate denoising window. === */
int2 low = make_int2(max(rect.x, x - radius),
max(rect.y, y - radius));
int2 high = make_int2(min(rect.z, x + radius + 1),
min(rect.w, y + radius + 1));
int num_pixels = (high.y - low.y) * (high.x - low.x);
/* === Shift feature passes to have mean 0. === */
float feature_means[DENOISE_FEATURES];
@ -52,8 +47,7 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
math_vector_add(feature_means, features, DENOISE_FEATURES);
} END_FOR_PIXEL_WINDOW
float pixel_scale = 1.0f / ((high.y - low.y) * (high.x - low.x));
math_vector_scale(feature_means, pixel_scale, DENOISE_FEATURES);
math_vector_scale(feature_means, 1.0f / num_pixels, DENOISE_FEATURES);
/* === Scale the shifted feature passes to a range of [-1; 1], will be baked into the transform later. === */
float *feature_scale = tempvector;
@ -66,7 +60,6 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
filter_calculate_scale(feature_scale);
/* === Generate the feature transformation. ===
* This transformation maps the DENOISE_FEATURES-dimentional feature space to a reduced feature (r-feature) space
* which generally has fewer dimensions. This mainly helps to prevent overfitting. */
@ -80,6 +73,8 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
math_matrix_jacobi_eigendecomposition(feature_matrix, transform, DENOISE_FEATURES, 1);
*rank = 0;
/* Prevent overfitting when a small window is used. */
int max_rank = min(DENOISE_FEATURES, num_pixels/3);
if(pca_threshold < 0.0f) {
float threshold_energy = 0.0f;
for(int i = 0; i < DENOISE_FEATURES; i++) {
@ -88,7 +83,7 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
threshold_energy *= 1.0f - (-pca_threshold);
float reduced_energy = 0.0f;
for(int i = 0; i < DENOISE_FEATURES; i++, (*rank)++) {
for(int i = 0; i < max_rank; i++, (*rank)++) {
if(i >= 2 && reduced_energy >= threshold_energy)
break;
float s = feature_matrix[i*DENOISE_FEATURES+i];
@ -96,7 +91,7 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
}
}
else {
for(int i = 0; i < DENOISE_FEATURES; i++, (*rank)++) {
for(int i = 0; i < max_rank; i++, (*rank)++) {
float s = feature_matrix[i*DENOISE_FEATURES+i];
if(i >= 2 && sqrtf(s) < pca_threshold)
break;

14
intern/cycles/kernel/filter/filter_transform_gpu.h
View File

@ -16,7 +16,7 @@
CCL_NAMESPACE_BEGIN
ccl_device void kernel_filter_construct_transform(ccl_global float ccl_restrict_ptr buffer,
ccl_device void kernel_filter_construct_transform(const ccl_global float *ccl_restrict buffer,
int x, int y, int4 rect,
int pass_stride,
ccl_global float *transform,
@ -38,7 +38,8 @@ ccl_device void kernel_filter_construct_transform(ccl_global float ccl_restrict_
max(rect.y, y - radius));
int2 high = make_int2(min(rect.z, x + radius + 1),
min(rect.w, y + radius + 1));
ccl_global float ccl_restrict_ptr pixel_buffer;
int num_pixels = (high.y - low.y) * (high.x - low.x);
const ccl_global float *ccl_restrict pixel_buffer;
int2 pixel;
@ -52,8 +53,7 @@ ccl_device void kernel_filter_construct_transform(ccl_global float ccl_restrict_
math_vector_add(feature_means, features, DENOISE_FEATURES);
} END_FOR_PIXEL_WINDOW
float pixel_scale = 1.0f / ((high.y - low.y) * (high.x - low.x));
math_vector_scale(feature_means, pixel_scale, DENOISE_FEATURES);
math_vector_scale(feature_means, 1.0f / num_pixels, DENOISE_FEATURES);
/* === Scale the shifted feature passes to a range of [-1; 1], will be baked into the transform later. === */
float feature_scale[DENOISE_FEATURES];
@ -81,6 +81,8 @@ ccl_device void kernel_filter_construct_transform(ccl_global float ccl_restrict_
math_matrix_jacobi_eigendecomposition(feature_matrix, transform, DENOISE_FEATURES, transform_stride);
*rank = 0;
/* Prevent overfitting when a small window is used. */
int max_rank = min(DENOISE_FEATURES, num_pixels/3);
if(pca_threshold < 0.0f) {
float threshold_energy = 0.0f;
for(int i = 0; i < DENOISE_FEATURES; i++) {
@ -89,7 +91,7 @@ ccl_device void kernel_filter_construct_transform(ccl_global float ccl_restrict_
threshold_energy *= 1.0f - (-pca_threshold);
float reduced_energy = 0.0f;
for(int i = 0; i < DENOISE_FEATURES; i++, (*rank)++) {
for(int i = 0; i < max_rank; i++, (*rank)++) {
if(i >= 2 && reduced_energy >= threshold_energy)
break;
float s = feature_matrix[i*DENOISE_FEATURES+i];
@ -97,7 +99,7 @@ ccl_device void kernel_filter_construct_transform(ccl_global float ccl_restrict_
}
}
else {
for(int i = 0; i < DENOISE_FEATURES; i++, (*rank)++) {
for(int i = 0; i < max_rank; i++, (*rank)++) {
float s = feature_matrix[i*DENOISE_FEATURES+i];
if(i >= 2 && sqrtf(s) < pca_threshold)
break;

13
intern/cycles/kernel/filter/filter_transform_sse.h
View File

@ -16,7 +16,7 @@
CCL_NAMESPACE_BEGIN
ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
ccl_device void kernel_filter_construct_transform(const float *ccl_restrict buffer,
int x, int y, int4 rect,
int pass_stride,
float *transform, int *rank,
@ -25,13 +25,14 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
int buffer_w = align_up(rect.z - rect.x, 4);
__m128 features[DENOISE_FEATURES];
float ccl_restrict_ptr pixel_buffer;
const float *ccl_restrict pixel_buffer;
int2 pixel;
int2 low = make_int2(max(rect.x, x - radius),
max(rect.y, y - radius));
int2 high = make_int2(min(rect.z, x + radius + 1),
min(rect.w, y + radius + 1));
int num_pixels = (high.y - low.y) * (high.x - low.x);
__m128 feature_means[DENOISE_FEATURES];
math_vector_zero_sse(feature_means, DENOISE_FEATURES);
@ -40,7 +41,7 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
math_vector_add_sse(feature_means, DENOISE_FEATURES, features);
} END_FOR_PIXEL_WINDOW_SSE
__m128 pixel_scale = _mm_set1_ps(1.0f / ((high.y - low.y) * (high.x - low.x)));
__m128 pixel_scale = _mm_set1_ps(1.0f / num_pixels);
for(int i = 0; i < DENOISE_FEATURES; i++) {
feature_means[i] = _mm_mul_ps(_mm_hsum_ps(feature_means[i]), pixel_scale);
}
@ -68,6 +69,8 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
math_matrix_jacobi_eigendecomposition(feature_matrix, transform, DENOISE_FEATURES, 1);
*rank = 0;
/* Prevent overfitting when a small window is used. */
int max_rank = min(DENOISE_FEATURES, num_pixels/3);
if(pca_threshold < 0.0f) {
float threshold_energy = 0.0f;
for(int i = 0; i < DENOISE_FEATURES; i++) {
@ -76,7 +79,7 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
threshold_energy *= 1.0f - (-pca_threshold);
float reduced_energy = 0.0f;
for(int i = 0; i < DENOISE_FEATURES; i++, (*rank)++) {
for(int i = 0; i < max_rank; i++, (*rank)++) {
if(i >= 2 && reduced_energy >= threshold_energy)
break;
float s = feature_matrix[i*DENOISE_FEATURES+i];
@ -84,7 +87,7 @@ ccl_device void kernel_filter_construct_transform(float ccl_restrict_ptr buffer,
}
}
else {
for(int i = 0; i < DENOISE_FEATURES; i++, (*rank)++) {
for(int i = 0; i < max_rank; i++, (*rank)++) {
float s = feature_matrix[i*DENOISE_FEATURES+i];
if(i >= 2 && sqrtf(s) < pca_threshold)
break;

2
intern/cycles/kernel/kernel_compat_cpu.h
View File

@ -42,8 +42,6 @@
#include "util/util_types.h"
#include "util/util_texture.h"
#define ccl_restrict_ptr const * __restrict
#define ccl_addr_space
#define ccl_local_id(d) 0

1
intern/cycles/kernel/kernel_compat_cuda.h
View File

@ -55,7 +55,6 @@
#define ccl_restrict __restrict__
#define ccl_align(n) __align__(n)
#define ccl_restrict_ptr const * __restrict__
#define CCL_MAX_LOCAL_SIZE (CUDA_THREADS_BLOCK_WIDTH*CUDA_THREADS_BLOCK_WIDTH)

2
intern/cycles/kernel/kernel_compat_opencl.h
View File

@ -50,8 +50,6 @@
# define ccl_addr_space
#endif
#define ccl_restrict_ptr const * __restrict__
#define ccl_local_id(d) get_local_id(d)
#define ccl_global_id(d) get_global_id(d)

24
intern/cycles/kernel/kernels/cpu/filter_cpu.h
View File

@ -72,40 +72,40 @@ void KERNEL_FUNCTION_FULL_NAME(filter_construct_transform)(float* buffer,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_difference)(int dx,
int dy,
float *weightImage,
float *weight_image,
float *variance,
float *differenceImage,
float *difference_image,
int* rect,
int w,
int channel_offset,
float a,
float k_2);
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_blur)(float *differenceImage,
float *outImage,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_blur)(float *difference_image,
float *out_image,
int* rect,
int w,
int f);
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_weight)(float *differenceImage,
float *outImage,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_weight)(float *difference_image,
float *out_image,
int* rect,
int w,
int f);
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_update_output)(int dx,
int dy,
float *differenceImage,
float *difference_image,
float *image,
float *outImage,
float *accumImage,
float *out_image,
float *accum_image,
int* rect,
int w,
int f);
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_construct_gramian)(int dx,
int dy,
float *differenceImage,
float *difference_image,
float *buffer,
float *color_pass,
float *variance_pass,
@ -120,8 +120,8 @@ void KERNEL_FUNCTION_FULL_NAME(filter_nlm_construct_gramian)(int dx,
int f,
int pass_stride);
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_normalize)(float *outImage,
float *accumImage,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_normalize)(float *out_image,
float *accum_image,
int* rect,
int w);

36
intern/cycles/kernel/kernels/cpu/filter_cpu_impl.h
View File

@ -150,9 +150,9 @@ void KERNEL_FUNCTION_FULL_NAME(filter_construct_transform)(float* buffer,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_difference)(int dx,
int dy,
float *weightImage,
float *weight_image,
float *variance,
float *differenceImage,
float *difference_image,
int *rect,
int w,
int channel_offset,
@ -162,12 +162,12 @@ void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_difference)(int dx,
#ifdef KERNEL_STUB
STUB_ASSERT(KERNEL_ARCH, filter_nlm_calc_difference);
#else
kernel_filter_nlm_calc_difference(dx, dy, weightImage, variance, differenceImage, load_int4(rect), w, channel_offset, a, k_2);
kernel_filter_nlm_calc_difference(dx, dy, weight_image, variance, difference_image, load_int4(rect), w, channel_offset, a, k_2);
#endif
}
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_blur)(float *differenceImage,
float *outImage,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_blur)(float *difference_image,
float *out_image,
int *rect,
int w,
int f)
@ -175,12 +175,12 @@ void KERNEL_FUNCTION_FULL_NAME(filter_nlm_blur)(float *differenceImage,
#ifdef KERNEL_STUB
STUB_ASSERT(KERNEL_ARCH, filter_nlm_blur);
#else
kernel_filter_nlm_blur(differenceImage, outImage, load_int4(rect), w, f);
kernel_filter_nlm_blur(difference_image, out_image, load_int4(rect), w, f);
#endif
}
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_weight)(float *differenceImage,
float *outImage,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_weight)(float *difference_image,
float *out_image,
int *rect,
int w,
int f)
@ -188,16 +188,16 @@ void KERNEL_FUNCTION_FULL_NAME(filter_nlm_calc_weight)(float *differenceImage,
#ifdef KERNEL_STUB
STUB_ASSERT(KERNEL_ARCH, filter_nlm_calc_weight);
#else
kernel_filter_nlm_calc_weight(differenceImage, outImage, load_int4(rect), w, f);
kernel_filter_nlm_calc_weight(difference_image, out_image, load_int4(rect), w, f);
#endif
}
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_update_output)(int dx,
int dy,
float *differenceImage,
float *difference_image,
float *image,
float *outImage,
float *accumImage,
float *out_image,
float *accum_image,
int *rect,
int w,
int f)
@ -205,13 +205,13 @@ void KERNEL_FUNCTION_FULL_NAME(filter_nlm_update_output)(int dx,
#ifdef KERNEL_STUB
STUB_ASSERT(KERNEL_ARCH, filter_nlm_update_output);
#else
kernel_filter_nlm_update_output(dx, dy, differenceImage, image, outImage, accumImage, load_int4(rect), w, f);
kernel_filter_nlm_update_output(dx, dy, difference_image, image, out_image, accum_image, load_int4(rect), w, f);
#endif
}
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_construct_gramian)(int dx,
int dy,
float *differenceImage,
float *difference_image,
float *buffer,
float *color_pass,
float *variance_pass,
@ -229,19 +229,19 @@ void KERNEL_FUNCTION_FULL_NAME(filter_nlm_construct_gramian)(int dx,
#ifdef KERNEL_STUB
STUB_ASSERT(KERNEL_ARCH, filter_nlm_construct_gramian);
#else
kernel_filter_nlm_construct_gramian(dx, dy, differenceImage, buffer, color_pass, variance_pass, transform, rank, XtWX, XtWY, load_int4(rect), load_int4(filter_rect), w, h, f, pass_stride);
kernel_filter_nlm_construct_gramian(dx, dy, difference_image, buffer, color_pass, variance_pass, transform, rank, XtWX, XtWY, load_int4(rect), load_int4(filter_rect), w, h, f, pass_stride);
#endif
}
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_normalize)(float *outImage,
float *accumImage,
void KERNEL_FUNCTION_FULL_NAME(filter_nlm_normalize)(float *out_image,
float *accum_image,
int *rect,
int w)
{
#ifdef KERNEL_STUB
STUB_ASSERT(KERNEL_ARCH, filter_nlm_normalize);
#else
kernel_filter_nlm_normalize(outImage, accumImage, load_int4(rect), w);
kernel_filter_nlm_normalize(out_image, accum_image, load_int4(rect), w);
#endif
}

49
intern/cycles/kernel/kernels/cuda/filter.cu
View File

@ -139,69 +139,74 @@ kernel_cuda_filter_construct_transform(float const* __restrict__ buffer,
extern "C" __global__ void
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
kernel_cuda_filter_nlm_calc_difference(int dx, int dy,
float ccl_restrict_ptr weightImage,
float ccl_restrict_ptr varianceImage,
float *differenceImage,
const float *ccl_restrict weight_image,
const float *ccl_restrict variance_image,
float *difference_image,
int4 rect, int w,
int channel_offset,
float a, float k_2) {
float a, float k_2)
{
int x = blockDim.x*blockIdx.x + threadIdx.x + rect.x;
int y = blockDim.y*blockIdx.y + threadIdx.y + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_calc_difference(x, y, dx, dy, weightImage, varianceImage, differenceImage, rect, w, channel_offset, a, k_2);
kernel_filter_nlm_calc_difference(x, y, dx, dy, weight_image, variance_image, difference_image, rect, w, channel_offset, a, k_2);
}
}
extern "C" __global__ void
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
kernel_cuda_filter_nlm_blur(float ccl_restrict_ptr differenceImage, float *outImage, int4 rect, int w, int f) {
kernel_cuda_filter_nlm_blur(const float *ccl_restrict difference_image, float *out_image, int4 rect, int w, int f)
{
int x = blockDim.x*blockIdx.x + threadIdx.x + rect.x;
int y = blockDim.y*blockIdx.y + threadIdx.y + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_blur(x, y, differenceImage, outImage, rect, w, f);
kernel_filter_nlm_blur(x, y, difference_image, out_image, rect, w, f);
}
}
extern "C" __global__ void
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
kernel_cuda_filter_nlm_calc_weight(float ccl_restrict_ptr differenceImage, float *outImage, int4 rect, int w, int f) {
kernel_cuda_filter_nlm_calc_weight(const float *ccl_restrict difference_image, float *out_image, int4 rect, int w, int f)
{
int x = blockDim.x*blockIdx.x + threadIdx.x + rect.x;
int y = blockDim.y*blockIdx.y + threadIdx.y + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_calc_weight(x, y, differenceImage, outImage, rect, w, f);
kernel_filter_nlm_calc_weight(x, y, difference_image, out_image, rect, w, f);
}
}
extern "C" __global__ void
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
kernel_cuda_filter_nlm_update_output(int dx, int dy,
float ccl_restrict_ptr differenceImage,
float ccl_restrict_ptr image,
float *outImage, float *accumImage,
const float *ccl_restrict difference_image,
const float *ccl_restrict image,
float *out_image, float *accum_image,
int4 rect, int w,
int f) {
int f)
{
int x = blockDim.x*blockIdx.x + threadIdx.x + rect.x;
int y = blockDim.y*blockIdx.y + threadIdx.y + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_update_output(x, y, dx, dy, differenceImage, image, outImage, accumImage, rect, w, f);
kernel_filter_nlm_update_output(x, y, dx, dy, difference_image, image, out_image, accum_image, rect, w, f);
}
}
extern "C" __global__ void
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
kernel_cuda_filter_nlm_normalize(float *outImage, float ccl_restrict_ptr accumImage, int4 rect, int w) {
kernel_cuda_filter_nlm_normalize(float *out_image, const float *ccl_restrict accum_image, int4 rect, int w)
{
int x = blockDim.x*blockIdx.x + threadIdx.x + rect.x;
int y = blockDim.y*blockIdx.y + threadIdx.y + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_normalize(x, y, outImage, accumImage, rect, w);
kernel_filter_nlm_normalize(x, y, out_image, accum_image, rect, w);
}
}
extern "C" __global__ void
CUDA_LAUNCH_BOUNDS(CUDA_THREADS_BLOCK_WIDTH, CUDA_KERNEL_MAX_REGISTERS)
kernel_cuda_filter_nlm_construct_gramian(int dx, int dy,
float ccl_restrict_ptr differenceImage,
float ccl_restrict_ptr buffer,
const float *ccl_restrict difference_image,
const float *ccl_restrict buffer,
float *color_pass,
float *variance_pass,
float const* __restrict__ transform,
@ -211,13 +216,14 @@ kernel_cuda_filter_nlm_construct_gramian(int dx, int dy,
int4 rect,
int4 filter_rect,
int w, int h, int f,
int pass_stride) {
int pass_stride)
{
int x = blockDim.x*blockIdx.x + threadIdx.x + max(0, rect.x-filter_rect.x);
int y = blockDim.y*blockIdx.y + threadIdx.y + max(0, rect.y-filter_rect.y);
if(x < min(filter_rect.z, rect.z-filter_rect.x) && y < min(filter_rect.w, rect.w-filter_rect.y)) {
kernel_filter_nlm_construct_gramian(x, y,
dx, dy,
differenceImage,
difference_image,
buffer,
color_pass, variance_pass,
transform, rank,
@ -235,7 +241,8 @@ kernel_cuda_filter_finalize(int w, int h,
float *buffer, int *rank,
float *XtWX, float3 *XtWY,
int4 filter_area, int4 buffer_params,
int sample) {
int sample)
{
int x = blockDim.x*blockIdx.x + threadIdx.x;
int y = blockDim.y*blockIdx.y + threadIdx.y;
if(x < filter_area.z && y < filter_area.w) {

67
intern/cycles/kernel/kernels/opencl/filter.cl
View File

@ -106,7 +106,7 @@ __kernel void kernel_ocl_filter_combine_halves(ccl_global float *mean,
}
}
__kernel void kernel_ocl_filter_construct_transform(ccl_global float ccl_restrict_ptr buffer,
__kernel void kernel_ocl_filter_construct_transform(const ccl_global float *ccl_restrict buffer,
ccl_global float *transform,
ccl_global int *rank,
int4 filter_area,
@ -132,79 +132,84 @@ __kernel void kernel_ocl_filter_construct_transform(ccl_global float ccl_restric
__kernel void kernel_ocl_filter_nlm_calc_difference(int dx,
int dy,
ccl_global float ccl_restrict_ptr weightImage,
ccl_global float ccl_restrict_ptr varianceImage,
ccl_global float *differenceImage,
const ccl_global float *ccl_restrict weight_image,
const ccl_global float *ccl_restrict variance_image,
ccl_global float *difference_image,
int4 rect,
int w,
int channel_offset,
float a,
float k_2) {
float k_2)
{
int x = get_global_id(0) + rect.x;
int y = get_global_id(1) + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_calc_difference(x, y, dx, dy, weightImage, varianceImage, differenceImage, rect, w, channel_offset, a, k_2);
kernel_filter_nlm_calc_difference(x, y, dx, dy, weight_image, variance_image, difference_image, rect, w, channel_offset, a, k_2);
}
}
__kernel void kernel_ocl_filter_nlm_blur(ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float *outImage,
__kernel void kernel_ocl_filter_nlm_blur(const ccl_global float *ccl_restrict difference_image,
ccl_global float *out_image,
int4 rect,
int w,
int f) {
int f)
{
int x = get_global_id(0) + rect.x;
int y = get_global_id(1) + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_blur(x, y, differenceImage, outImage, rect, w, f);
kernel_filter_nlm_blur(x, y, difference_image, out_image, rect, w, f);
}
}
__kernel void kernel_ocl_filter_nlm_calc_weight(ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float *outImage,
__kernel void kernel_ocl_filter_nlm_calc_weight(const ccl_global float *ccl_restrict difference_image,
ccl_global float *out_image,
int4 rect,
int w,
int f) {
int f)
{
int x = get_global_id(0) + rect.x;
int y = get_global_id(1) + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_calc_weight(x, y, differenceImage, outImage, rect, w, f);
kernel_filter_nlm_calc_weight(x, y, difference_image, out_image, rect, w, f);
}
}
__kernel void kernel_ocl_filter_nlm_update_output(int dx,
int dy,
ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float ccl_restrict_ptr image,
ccl_global float *outImage,
ccl_global float *accumImage,
const ccl_global float *ccl_restrict difference_image,
const ccl_global float *ccl_restrict image,
ccl_global float *out_image,
ccl_global float *accum_image,
int4 rect,
int w,
int f) {
int f)
{
int x = get_global_id(0) + rect.x;
int y = get_global_id(1) + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_update_output(x, y, dx, dy, differenceImage, image, outImage, accumImage, rect, w, f);
kernel_filter_nlm_update_output(x, y, dx, dy, difference_image, image, out_image, accum_image, rect, w, f);
}
}
__kernel void kernel_ocl_filter_nlm_normalize(ccl_global float *outImage,
ccl_global float ccl_restrict_ptr accumImage,
__kernel void kernel_ocl_filter_nlm_normalize(ccl_global float *out_image,
const ccl_global float *ccl_restrict accum_image,
int4 rect,
int w) {
int w)
{
int x = get_global_id(0) + rect.x;
int y = get_global_id(1) + rect.y;
if(x < rect.z && y < rect.w) {
kernel_filter_nlm_normalize(x, y, outImage, accumImage, rect, w);
kernel_filter_nlm_normalize(x, y, out_image, accum_image, rect, w);
}
}
__kernel void kernel_ocl_filter_nlm_construct_gramian(int dx,
int dy,
ccl_global float ccl_restrict_ptr differenceImage,
ccl_global float ccl_restrict_ptr buffer,
const ccl_global float *ccl_restrict difference_image,
const ccl_global float *ccl_restrict buffer,
ccl_global float *color_pass,
ccl_global float *variance_pass,
ccl_global float ccl_restrict_ptr transform,
const ccl_global float *ccl_restrict transform,
ccl_global int *rank,
ccl_global float *XtWX,
ccl_global float3 *XtWY,
@ -213,13 +218,14 @@ __kernel void kernel_ocl_filter_nlm_construct_gramian(int dx,
int w,
int h,
int f,
int pass_stride) {
int pass_stride)
{
int x = get_global_id(0) + max(0, rect.x-filter_rect.x);
int y = get_global_id(1) + max(0, rect.y-filter_rect.y);
if(x < min(filter_rect.z, rect.z-filter_rect.x) && y < min(filter_rect.w, rect.w-filter_rect.y)) {
kernel_filter_nlm_construct_gramian(x, y,
dx, dy,
differenceImage,
difference_image,
buffer,
color_pass, variance_pass,
transform, rank,
@ -239,7 +245,8 @@ __kernel void kernel_ocl_filter_finalize(int w,
ccl_global float3 *XtWY,
int4 filter_area,
int4 buffer_params,
int sample) {
int sample)
{
int x = get_global_id(0);
int y = get_global_id(1);
if(x < filter_area.z && y < filter_area.w) {

1
intern/cycles/kernel/split/kernel_buffer_update.h
View File

@ -111,7 +111,6 @@ ccl_device void kernel_buffer_update(KernelGlobals *kg,
buffer += (kernel_split_params.offset + pixel_x + pixel_y*stride) * kernel_data.film.pass_stride;
if(IS_STATE(ray_state, ray_index, RAY_UPDATE_BUFFER)) {
kernel_write_light_passes(kg, buffer, L, sample);
#ifdef __KERNEL_DEBUG__
kernel_write_debug_passes(kg, buffer, state, debug_data, sample);
#endif

15
intern/cycles/render/session.cpp
View File

@ -726,16 +726,20 @@ DeviceRequestedFeatures Session::get_requested_device_features()
return requested_features;
}
void Session::load_kernels()
void Session::load_kernels(bool lock_scene)
{
thread_scoped_lock scene_lock(scene->mutex);
thread_scoped_lock scene_lock;
if(lock_scene) {
scene_lock = thread_scoped_lock(scene->mutex);
}
if(!kernels_loaded) {
DeviceRequestedFeatures requested_features = get_requested_device_features();
if(!kernels_loaded || loaded_kernel_features.modified(requested_features)) {
progress.set_status("Loading render kernels (may take a few minutes the first time)");
scoped_timer timer;
DeviceRequestedFeatures requested_features = get_requested_device_features();
VLOG(2) << "Requested features:\n" << requested_features;
if(!device->load_kernels(requested_features)) {
string message = device->error_message();
@ -752,6 +756,7 @@ void Session::load_kernels()
VLOG(1) << "Total time spent loading kernels: " << time_dt() - timer.get_start();
kernels_loaded = true;
loaded_kernel_features = requested_features;
}
}
@ -902,6 +907,8 @@ void Session::update_scene()
/* update scene */
if(scene->need_update()) {
load_kernels(false);
progress.set_status("Updating Scene");
MEM_GUARDED_CALL(&progress, scene->device_update, device, progress);
}

3
intern/cycles/render/session.h
View File

@ -158,7 +158,7 @@ public:
void set_pause(bool pause);
void update_scene();
void load_kernels();
void load_kernels(bool lock_scene=true);
void device_free();
@ -215,6 +215,7 @@ protected:
thread_mutex display_mutex;
bool kernels_loaded;
DeviceRequestedFeatures loaded_kernel_features;
double reset_time;

5
intern/cycles/util/util_math_float3.h
View File

@ -367,6 +367,11 @@ ccl_device_inline bool isequal_float3(const float3 a, const float3 b)
#endif
}
ccl_device_inline bool isfinite3_safe(float3 v)
{
return isfinite_safe(v.x) && isfinite_safe(v.y) && isfinite_safe(v.z);
}
ccl_device_inline float3 ensure_finite3(float3 v)
{
if(!isfinite_safe(v.x)) v.x = 0.0;

130
intern/cycles/util/util_math_matrix.h
View File

@ -23,73 +23,83 @@ CCL_NAMESPACE_BEGIN
/* Variants that use a constant stride on GPUS. */
#ifdef __KERNEL_GPU__
#define MATS(A, n, r, c, s) A[((r)*(n)+(c))*(s)]
# define MATS(A, n, r, c, s) A[((r)*(n)+(c))*(s)]
/* Element access when only the lower-triangular elements are stored. */
#define MATHS(A, r, c, s) A[((r)*((r)+1)/2+(c))*(s)]
#define VECS(V, i, s) V[(i)*(s)]
# define MATHS(A, r, c, s) A[((r)*((r)+1)/2+(c))*(s)]
# define VECS(V, i, s) V[(i)*(s)]
#else
#define MATS(A, n, r, c, s) MAT(A, n, r, c)
#define MATHS(A, r, c, s) A[(r)*((r)+1)/2+(c)]
#define VECS(V, i, s) V[i]
# define MATS(A, n, r, c, s) MAT(A, n, r, c)
# define MATHS(A, r, c, s) A[(r)*((r)+1)/2+(c)]
# define VECS(V, i, s) V[i]
#endif
/* Zeroing helpers. */
ccl_device_inline void math_vector_zero(float *v, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
v[i] = 0.0f;
}
}
ccl_device_inline void math_matrix_zero(float *A, int n)
{
for(int row = 0; row < n; row++)
for(int col = 0; col <= row; col++)
for(int row = 0; row < n; row++) {
for(int col = 0; col <= row; col++) {
MAT(A, n, row, col) = 0.0f;
}
}
}
/* Elementary vector operations. */
ccl_device_inline void math_vector_add(float *a, float ccl_restrict_ptr b, int n)
ccl_device_inline void math_vector_add(float *a, const float *ccl_restrict b, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
a[i] += b[i];
}
}
ccl_device_inline void math_vector_mul(float *a, float ccl_restrict_ptr b, int n)
ccl_device_inline void math_vector_mul(float *a, const float *ccl_restrict b, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
a[i] *= b[i];
}
}
ccl_device_inline void math_vector_mul_strided(ccl_global float *a, float ccl_restrict_ptr b, int astride, int n)
ccl_device_inline void math_vector_mul_strided(ccl_global float *a, const float *ccl_restrict b, int astride, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
a[i*astride] *= b[i];
}
}
ccl_device_inline void math_vector_scale(float *a, float b, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
a[i] *= b;
}
}
ccl_device_inline void math_vector_max(float *a, float ccl_restrict_ptr b, int n)
ccl_device_inline void math_vector_max(float *a, const float *ccl_restrict b, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
a[i] = max(a[i], b[i]);
}
}
ccl_device_inline void math_vec3_add(float3 *v, int n, float *x, float3 w)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
v[i] += w*x[i];
}
}
ccl_device_inline void math_vec3_add_strided(ccl_global float3 *v, int n, float *x, float3 w, int stride)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
v[i*stride] += w*x[i];
}
}
/* Elementary matrix operations.
@ -97,33 +107,38 @@ ccl_device_inline void math_vec3_add_strided(ccl_global float3 *v, int n, float
ccl_device_inline void math_trimatrix_add_diagonal(ccl_global float *A, int n, float val, int stride)
{
for(int row = 0; row < n; row++)
for(int row = 0; row < n; row++) {
MATHS(A, row, row, stride) += val;
}
}
/* Add Gramian matrix of v to A.
* The Gramian matrix of v is vt*v, so element (i,j) is v[i]*v[j]. */
ccl_device_inline void math_matrix_add_gramian(float *A,
int n,
float ccl_restrict_ptr v,
const float *ccl_restrict v,
float weight)
{
for(int row = 0; row < n; row++)
for(int col = 0; col <= row; col++)
for(int row = 0; row < n; row++) {
for(int col = 0; col <= row; col++) {
MAT(A, n, row, col) += v[row]*v[col]*weight;
}
}
}
/* Add Gramian matrix of v to A.
* The Gramian matrix of v is vt*v, so element (i,j) is v[i]*v[j]. */
ccl_device_inline void math_trimatrix_add_gramian_strided(ccl_global float *A,
int n,
float ccl_restrict_ptr v,
const float *ccl_restrict v,
float weight,
int stride)
{
for(int row = 0; row < n; row++)
for(int col = 0; col <= row; col++)
for(int row = 0; row < n; row++) {
for(int col = 0; col <= row; col++) {
MATHS(A, row, col, stride) += v[row]*v[col]*weight;
}
}
}
/* Transpose matrix A inplace. */
@ -138,9 +153,6 @@ ccl_device_inline void math_matrix_transpose(ccl_global float *A, int n, int str
}
}
/* Solvers for matrix problems */
/* In-place Cholesky-Banachiewicz decomposition of the square, positive-definite matrix A
@ -199,10 +211,6 @@ ccl_device_inline void math_trimatrix_vec3_solve(ccl_global float *A, ccl_global
}
}
/* Perform the Jacobi Eigenvalue Methon on matrix A.
* A is assumed to be a symmetrical matrix, therefore only the lower-triangular part is ever accessed.
* The algorithm overwrites the contents of A.
@ -215,15 +223,19 @@ ccl_device void math_matrix_jacobi_eigendecomposition(float *A, ccl_global float
{
const float singular_epsilon = 1e-9f;
for (int row = 0; row < n; row++)
for (int col = 0; col < n; col++)
for (int row = 0; row < n; row++) {
for (int col = 0; col < n; col++) {
MATS(V, n, row, col, v_stride) = (col == row) ? 1.0f : 0.0f;
}
}
for (int sweep = 0; sweep < 8; sweep++) {
float off_diagonal = 0.0f;
for (int row = 1; row < n; row++)
for (int col = 0; col < row; col++)
for (int row = 1; row < n; row++) {
for (int col = 0; col < row; col++) {
off_diagonal += fabsf(MAT(A, n, row, col));
}
}
if (off_diagonal < 1e-7f) {
/* The matrix has nearly reached diagonal form.
* Since the eigenvalues are only used to determine truncation, their exact values aren't required - a relative error of a few ULPs won't matter at all. */
@ -327,51 +339,61 @@ ccl_device void math_matrix_jacobi_eigendecomposition(float *A, ccl_global float
}
#ifdef __KERNEL_SSE3__
ccl_device_inline void math_vector_zero_sse(__m128 *A, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
A[i] = _mm_setzero_ps();
}
}
ccl_device_inline void math_matrix_zero_sse(__m128 *A, int n)
{
for(int row = 0; row < n; row++)
for(int col = 0; col <= row; col++)
for(int row = 0; row < n; row++) {
for(int col = 0; col <= row; col++) {
MAT(A, n, row, col) = _mm_setzero_ps();
}
}
}
/* Add Gramian matrix of v to A.
* The Gramian matrix of v is v^T*v, so element (i,j) is v[i]*v[j]. */
ccl_device_inline void math_matrix_add_gramian_sse(__m128 *A, int n, __m128 ccl_restrict_ptr v, __m128 weight)
ccl_device_inline void math_matrix_add_gramian_sse(__m128 *A, int n, const __m128 *ccl_restrict v, __m128 weight)
{
for(int row = 0; row < n; row++)
for(int col = 0; col <= row; col++)
for(int row = 0; row < n; row++) {
for(int col = 0; col <= row; col++) {
MAT(A, n, row, col) = _mm_add_ps(MAT(A, n, row, col), _mm_mul_ps(_mm_mul_ps(v[row], v[col]), weight));
}
}
}
ccl_device_inline void math_vector_add_sse(__m128 *V, int n, __m128 ccl_restrict_ptr a)
ccl_device_inline void math_vector_add_sse(__m128 *V, int n, const __m128 *ccl_restrict a)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
V[i] = _mm_add_ps(V[i], a[i]);
}
}
ccl_device_inline void math_vector_mul_sse(__m128 *V, int n, __m128 ccl_restrict_ptr a)
ccl_device_inline void math_vector_mul_sse(__m128 *V, int n, const __m128 *ccl_restrict a)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
V[i] = _mm_mul_ps(V[i], a[i]);
}
}
ccl_device_inline void math_vector_max_sse(__m128 *a, __m128 ccl_restrict_ptr b, int n)
ccl_device_inline void math_vector_max_sse(__m128 *a, const __m128 *ccl_restrict b, int n)
{
for(int i = 0; i < n; i++)
for(int i = 0; i < n; i++) {
a[i] = _mm_max_ps(a[i], b[i]);
}
}
ccl_device_inline void math_matrix_hsum(float *A, int n, __m128 ccl_restrict_ptr B)
ccl_device_inline void math_matrix_hsum(float *A, int n, const __m128 *ccl_restrict B)
{
for(int row = 0; row < n; row++)
for(int col = 0; col <= row; col++)
for(int row = 0; row < n; row++) {
for(int col = 0; col <= row; col++) {
MAT(A, n, row, col) = _mm_hsum_ss(MAT(B, n, row, col));
}
}
}
#endif

28
intern/string/STR_HashedString.h
View File

@ -38,6 +38,14 @@
#include "STR_String.h"
/* copied from 'BLI_compiler_attrs.h' */
/* Use to suppress '-Wimplicit-fallthrough' (in place of 'break'). */
#if defined(__GNUC__) && (__GNUC__ >= 7) /* gcc7.0+ only */
#define ATTR_FALLTHROUGH __attribute__((fallthrough))
#else
#define ATTR_FALLTHROUGH ((void)0)
#endif
// Hash Mix utility function, by Bob Jenkins - Mix 3 32-bit values reversibly
//
@ -102,16 +110,16 @@ static dword STR_gHash(const void *in, int len, dword init_val)
// Handle the last 11 bytes
c += len;
switch (length) {
case 11: c += ((dword)p_in[10] << 24);
case 10: c += ((dword)p_in[9] << 16);
case 9: c += ((dword)p_in[8] << 8); /* the first byte of c is reserved for the length */
case 8: b += ((dword)p_in[7] << 24);
case 7: b += ((dword)p_in[6] << 16);
case 6: b += ((dword)p_in[5] << 8);
case 5: b += p_in[4];
case 4: a += ((dword)p_in[3] << 24);
case 3: a += ((dword)p_in[2] << 16);
case 2: a += ((dword)p_in[1] << 8);
case 11: c += ((dword)p_in[10] << 24); ATTR_FALLTHROUGH;
case 10: c += ((dword)p_in[9] << 16); ATTR_FALLTHROUGH;
case 9: c += ((dword)p_in[8] << 8); ATTR_FALLTHROUGH; /* the first byte of c is reserved for the length */
case 8: b += ((dword)p_in[7] << 24); ATTR_FALLTHROUGH;
case 7: b += ((dword)p_in[6] << 16); ATTR_FALLTHROUGH;
case 6: b += ((dword)p_in[5] << 8); ATTR_FALLTHROUGH;
case 5: b += p_in[4]; ATTR_FALLTHROUGH;
case 4: a += ((dword)p_in[3] << 24); ATTR_FALLTHROUGH;
case 3: a += ((dword)p_in[2] << 16); ATTR_FALLTHROUGH;
case 2: a += ((dword)p_in[1] << 8); ATTR_FALLTHROUGH;
case 1: a += p_in[0];
}
STR_gHashMix(a, b, c);

1
source/blender/alembic/intern/abc_curves.cc
View File

@ -297,6 +297,7 @@ void read_curve_sample(Curve *cu, const ICurvesSchema &schema, const ISampleSele
nu->orderu = static_cast<short>((*orders)[i]);
break;
}
ATTR_FALLTHROUGH;
case Alembic::AbcGeom::kLinear:
default:
nu->orderu = 2;

2
source/blender/blenkernel/intern/image.c
View File

@ -1238,7 +1238,7 @@ char BKE_imtype_valid_channels(const char imtype, bool write_file)
switch (imtype) {
case R_IMF_IMTYPE_BMP:
if (write_file) break;
/* fall-through */
ATTR_FALLTHROUGH;
case R_IMF_IMTYPE_TARGA:
case R_IMF_IMTYPE_RAWTGA:
case R_IMF_IMTYPE_IRIS:

11
source/blender/blenkernel/intern/mball_tessellate.c
View File

@ -317,12 +317,12 @@ static float densfunc(const MetaElem *ball, float x, float y, float z)
if (dvec[2] > ball->expz) dvec[2] -= ball->expz;
else if (dvec[2] < -ball->expz) dvec[2] += ball->expz;
else dvec[2] = 0.0;
/* fall through */
ATTR_FALLTHROUGH;
case MB_PLANE:
if (dvec[1] > ball->expy) dvec[1] -= ball->expy;
else if (dvec[1] < -ball->expy) dvec[1] += ball->expy;
else dvec[1] = 0.0;
/* fall through */
ATTR_FALLTHROUGH;
case MB_TUBE:
if (dvec[0] > ball->expx) dvec[0] -= ball->expx;
else if (dvec[0] < -ball->expx) dvec[0] += ball->expx;
@ -661,7 +661,7 @@ static void makecubetable(void)
for (i = 0; i < 256; i++) {
for (e = 0; e < 12; e++) done[e] = 0;
for (c = 0; c < 8; c++) pos[c] = MB_BIT(i, c);
for (e = 0; e < 12; e++)
for (e = 0; e < 12; e++) {
if (!done[e] && (pos[corner1[e]] != pos[corner2[e]])) {
INTLIST *ints = NULL;
INTLISTS *lists = MEM_callocN(sizeof(INTLISTS), "mball_intlist");
@ -688,6 +688,7 @@ static void makecubetable(void)
lists->next = cubetable[i];
cubetable[i] = lists;
}
}
}
for (i = 0; i < 256; i++) {
@ -1194,10 +1195,10 @@ static void init_meta(EvaluationContext *eval_ctx, PROCESS *process, Scene *scen
break;
case MB_CUBE: /* cube is "expanded" by expz, expy and expx */
expz += ml->expz;
/* fall through */
ATTR_FALLTHROUGH;
case MB_PLANE: /* plane is "expanded" by expy and expx */
expy += ml->expy;
/* fall through */
ATTR_FALLTHROUGH;
case MB_TUBE: /* tube is "expanded" by expx */
expx += ml->expx;
break;

4
source/blender/blenkernel/intern/packedFile.c
View File

@ -450,7 +450,7 @@ char *unpackFile(ReportList *reports, const char *abs_name, const char *local_na
break;
}
/* else create it */
/* fall-through */
ATTR_FALLTHROUGH;
}
case PF_WRITE_LOCAL:
if (writePackedFile(reports, local_name, pf, 1) == RET_OK) {
@ -471,7 +471,7 @@ char *unpackFile(ReportList *reports, const char *abs_name, const char *local_na
break;
}
/* else create it */
/* fall-through */
ATTR_FALLTHROUGH;
}
case PF_WRITE_ORIGINAL:
if (writePackedFile(reports, abs_name, pf, 1) == RET_OK) {

6
source/blender/blenkernel/intern/particle.c
View File

@ -3471,7 +3471,8 @@ static void get_cpa_texture(DerivedMesh *dm, ParticleSystem *psys, ParticleSetti
case TEXCO_UV:
if (fw && get_particle_uv(dm, NULL, face_index, fw, mtex->uvname, texvec))
break;
/* no break, failed to get uv's, so let's try orco's */
/* no break, failed to get uv's, so let's try orco's */
ATTR_FALLTHROUGH;
case TEXCO_ORCO:
copy_v3_v3(texvec, orco);
break;
@ -3543,7 +3544,8 @@ void psys_get_texture(ParticleSimulationData *sim, ParticleData *pa, ParticleTex
case TEXCO_UV:
if (get_particle_uv(sim->psmd->dm_final, pa, 0, pa->fuv, mtex->uvname, texvec))
break;
/* no break, failed to get uv's, so let's try orco's */
/* no break, failed to get uv's, so let's try orco's */
ATTR_FALLTHROUGH;
case TEXCO_ORCO:
psys_particle_on_emitter(sim->psmd, sim->psys->part->from, pa->num, pa->num_dmcache, pa->fuv, pa->foffset, co, 0, 0, 0, texvec, 0);

2
source/blender/blenkernel/intern/rigidbody.c
View File

@ -821,7 +821,7 @@ static void rigidbody_validate_sim_constraint(RigidBodyWorld *rbw, Object *ob, b
RB_constraint_set_damping_6dof_spring(rbc->physics_constraint, RB_LIMIT_ANG_Z, rbc->spring_damping_ang_z);
RB_constraint_set_equilibrium_6dof_spring(rbc->physics_constraint);
/* fall-through */
ATTR_FALLTHROUGH;
case RBC_TYPE_6DOF:
if (rbc->type == RBC_TYPE_6DOF) /* a litte awkward but avoids duplicate code for limits */
rbc->physics_constraint = RB_constraint_new_6dof(loc, rot, rb1, rb2);

15
source/blender/blenkernel/intern/seqmodifier.c
View File

@ -202,21 +202,22 @@ static void whiteBalance_apply_threaded(int width, int height, unsigned char *re
for (y = 0; y < height; y++) {
for (x = 0; x < width; x++) {
int pixel_index = (y * width + x) * 4;
float result[4], mask[3] = {1.0f, 1.0f, 1.0f};
float rgba[4], result[4], mask[3] = {1.0f, 1.0f, 1.0f};
if (rect_float) {
copy_v3_v3(result, rect_float + pixel_index);
copy_v3_v3(rgba, rect_float + pixel_index);
}
else {
straight_uchar_to_premul_float(result, rect + pixel_index);
straight_uchar_to_premul_float(rgba, rect + pixel_index);
}
copy_v4_v4(result, rgba);
#if 0
mul_v3_v3(result, multiplier);
#else
/* similar to division without the clipping */
for (int i = 0; i < 3; i++) {
result[i] = 1.0f - powf(1.0f - result[i], multiplier[i]);
result[i] = 1.0f - powf(1.0f - rgba[i], multiplier[i]);
}
#endif
@ -227,9 +228,9 @@ static void whiteBalance_apply_threaded(int width, int height, unsigned char *re
rgb_uchar_to_float(mask, mask_rect + pixel_index);
}
result[0] = result[0] * (1.0f - mask[0]) + result[0] * mask[0];
result[1] = result[1] * (1.0f - mask[1]) + result[1] * mask[1];
result[2] = result[2] * (1.0f - mask[2]) + result[2] * mask[2];
result[0] = rgba[0] * (1.0f - mask[0]) + result[0] * mask[0];
result[1] = rgba[1] * (1.0f - mask[1]) + result[1] * mask[1];
result[2] = rgba[2] * (1.0f - mask[2]) + result[2] * mask[2];
if (rect_float) {
copy_v3_v3(rect_float + pixel_index, result);

6
source/blender/blenkernel/intern/softbody.c
View File

@ -1856,19 +1856,21 @@ static void dfdx_spring(int ia, int ic, int op, float dir[3], float L, float len
float m, delta_ij;
int i, j;
if (L < len) {
for (i=0;i<3;i++)
for (i=0;i<3;i++) {
for (j=0;j<3;j++) {
delta_ij = (i==j ? (1.0f): (0.0f));
m=factor*(dir[i]*dir[j] + (1-L/len)*(delta_ij - dir[i]*dir[j]));
EIG_linear_solver_matrix_add(ia+i, op+ic+j, m);
}
}
}
else {
for (i=0;i<3;i++)
for (i=0;i<3;i++) {
for (j=0;j<3;j++) {
m=factor*dir[i]*dir[j];
EIG_linear_solver_matrix_add(ia+i, op+ic+j, m);
}
}
}
}

8
source/blender/blenlib/BLI_compiler_attrs.h
View File

@ -92,4 +92,12 @@
# define ATTR_PRINTF_FORMAT(format_param, dots_param)
#endif
/* Use to suppress '-Wimplicit-fallthrough' (in place of 'break'). */
#if defined(__GNUC__) && (__GNUC__ >= 7) /* gcc7.0+ only */
#define ATTR_FALLTHROUGH __attribute__((fallthrough))
#else
#define ATTR_FALLTHROUGH ((void)0)
#endif
#endif /* __BLI_COMPILER_ATTRS_H__ */

6
source/blender/blenlib/intern/hash_mm2a.c
View File

@ -36,6 +36,8 @@
* for temporary data.
*/
#include "BLI_compiler_attrs.h"
#include "BLI_hash_mm2a.h" /* own include */
/* Helpers. */
@ -128,10 +130,10 @@ uint32_t BLI_hash_mm2(const unsigned char *data, size_t len, uint32_t seed)
switch (len) {
case 3:
h ^= data[2] << 16;
/* fall through */
ATTR_FALLTHROUGH;
case 2:
h ^= data[1] << 8;
/* fall through */
ATTR_FALLTHROUGH;
case 1:
h ^= data[0];
h *= MM2A_M;

4
source/blender/blenlib/intern/string.c
View File

@ -332,7 +332,7 @@ size_t BLI_strescape(char *__restrict dst, const char *__restrict src, const siz
goto escape_finish;
case '\\':
case '"':
/* fall-through */
ATTR_FALLTHROUGH;
/* less common but should also be support */
case '\t':
@ -346,7 +346,7 @@ size_t BLI_strescape(char *__restrict dst, const char *__restrict src, const siz
/* not enough space to escape */
break;
}
/* fall-through */
ATTR_FALLTHROUGH;
default:
*dst = *src;
break;

10
source/blender/blenlib/intern/string_utf8.c
View File

@ -205,11 +205,11 @@ int BLI_utf8_invalid_strip(char *str, size_t length)
while (*src != '\0' && (utf8_size = utf8_skip_data[*src]) < maxncpy) {\
maxncpy -= utf8_size; \
switch (utf8_size) { \
case 6: *dst ++ = *src ++; \
case 5: *dst ++ = *src ++; \
case 4: *dst ++ = *src ++; \
case 3: *dst ++ = *src ++; \
case 2: *dst ++ = *src ++; \
case 6: *dst ++ = *src ++; ATTR_FALLTHROUGH; \
case 5: *dst ++ = *src ++; ATTR_FALLTHROUGH; \
case 4: *dst ++ = *src ++; ATTR_FALLTHROUGH; \
case 3: *dst ++ = *src ++; ATTR_FALLTHROUGH; \
case 2: *dst ++ = *src ++; ATTR_FALLTHROUGH; \
case 1: *dst ++ = *src ++; \
} \
} \

3
source/blender/blenloader/intern/readfile.c
View File

@ -8803,7 +8803,8 @@ BlendFileData *blo_read_file_internal(FileData *fd, const char *filepath)
/* in 2.50+ files, the file identifier for screens is patched, forward compatibility */
case ID_SCRN:
bhead->code = ID_SCR;
/* deliberate pass on to default */
/* pass on to default */
ATTR_FALLTHROUGH;
default:
if (fd->skip_flags & BLO_READ_SKIP_DATA) {
bhead = blo_nextbhead(fd, bhead);

3
source/blender/blenloader/intern/versioning_260.c
View File

@ -1000,7 +1000,7 @@ void blo_do_versions_260(FileData *fd, Library *UNUSED(lib), Main *main)
{
/* convert extended ascii to utf-8 for text editor */
Text *text;
for (text = main->text.first; text; text = text->id.next)
for (text = main->text.first; text; text = text->id.next) {
if (!(text->flags & TXT_ISEXT)) {
TextLine *tl;
@ -1013,6 +1013,7 @@ void blo_do_versions_260(FileData *fd, Library *UNUSED(lib), Main *main)
text->curc = 0;
}
}
}
}
{
/* set new dynamic paint values */

2
source/blender/blenloader/intern/versioning_270.c
View File

@ -1615,6 +1615,8 @@ void blo_do_versions_270(FileData *fd, Library *UNUSED(lib), Main *main)
bNode *node;
for (node = ntree->nodes.first; node; node = node->next) {
if (node->type == CMP_NODE_R_LAYERS) {
/* Make sure new sockets are properly created. */
node_verify_socket_templates(ntree, node);
int pass_index = 0;
const char *sockname;
for (bNodeSocket *sock = node->outputs.first; sock && pass_index < 31; sock = sock->next, pass_index++) {

3
source/blender/blenloader/intern/versioning_legacy.c
View File

@ -2296,11 +2296,12 @@ void blo_do_versions_pre250(FileData *fd, Library *lib, Main *main)
/* during 2.41 images with this name were used for viewer node output, lets fix that */
if (main->versionfile == 241) {
Image *ima;
for (ima = main->image.first; ima; ima = ima->id.next)
for (ima = main->image.first; ima; ima = ima->id.next) {
if (STREQ(ima->name, "Compositor")) {
strcpy(ima->id.name + 2, "Viewer Node");
strcpy(ima->name, "Viewer Node");
}
}
}
}

29
source/blender/compositor/operations/COM_TextureOperation.cpp
View File

@ -41,6 +41,7 @@ TextureBaseOperation::TextureBaseOperation() : NodeOperation()
this->m_rd = NULL;
this->m_pool = NULL;
this->m_sceneColorManage = false;
setComplex(true);
}
TextureOperation::TextureOperation() : TextureBaseOperation()
{
@ -155,31 +156,3 @@ void TextureBaseOperation::executePixelSampled(float output[4], float x, float y
output[0] = output[1] = output[2] = output[3];
}
}
MemoryBuffer *TextureBaseOperation::createMemoryBuffer(rcti * /*rect2*/)
{
int height = getHeight();
int width = getWidth();
DataType datatype = this->getOutputSocket()->getDataType();
int add = 4;
if (datatype == COM_DT_VALUE) {
add = 1;
}
rcti rect;
rect.xmin = 0;
rect.ymin = 0;
rect.xmax = width;
rect.ymax = height;
MemoryBuffer *result = new MemoryBuffer(datatype, &rect);
float *data = result->getBuffer();
for (int y = 0; y < height; y++) {
for (int x = 0; x < width; x++, data += add) {
this->executePixelSampled(data, x, y, COM_PS_NEAREST);
}
}
return result;
}

2
source/blender/compositor/operations/COM_TextureOperation.h
View File

@ -59,8 +59,6 @@ protected:
* Constructor
*/
TextureBaseOperation();
MemoryBuffer *createMemoryBuffer(rcti *rect2);
public:
void executePixelSampled(float output[4], float x, float y, PixelSampler sampler);

1
source/blender/editors/animation/anim_draw.c
View File

@ -601,6 +601,7 @@ void ANIM_center_frame(struct bContext *C, int smooth_viewtx)
break;
}
/* else drop through, keep range instead */
ATTR_FALLTHROUGH;
case ZOOM_FRAME_MODE_KEEP_RANGE:
default:

3
source/blender/editors/animation/anim_markers.c
View File

@ -906,7 +906,8 @@ static int ed_marker_move_modal(bContext *C, wmOperator *op, const wmEvent *even
ed_marker_move_cancel(C, op);
return OPERATOR_CANCELLED;
}
/* else continue; <--- see if release event should be caught for tweak-end */
/* else continue; <--- see if release event should be caught for tweak-end */
ATTR_FALLTHROUGH;
case RETKEY:
case PADENTER:

2
source/blender/editors/armature/pose_slide.c
View File

@ -885,6 +885,7 @@ static int pose_slide_modal(bContext *C, wmOperator *op, const wmEvent *event)
break;
}
default:
{
if ((event->val == KM_PRESS) && handleNumInput(C, &pso->num, event)) {
float value;
@ -972,6 +973,7 @@ static int pose_slide_modal(bContext *C, wmOperator *op, const wmEvent *event)
/* allow to pass through */
return OPERATOR_RUNNING_MODAL | OPERATOR_PASS_THROUGH;
}
}
}

2
source/blender/editors/armature/reeb.c
View File

@ -2240,9 +2240,9 @@ static void glueByMergeSort(ReebGraph *rg, ReebArc *a0, ReebArc *a1, ReebEdge *e
else {
a1 = nextArcMappedToEdge(a1, e1);
}
}
}
}
}
static void mergePaths(ReebGraph *rg, ReebEdge *e0, ReebEdge *e1, ReebEdge *e2)
{

2
source/blender/editors/gpencil/gpencil_interpolate.c
View File

@ -625,6 +625,7 @@ static int gpencil_interpolate_modal(bContext *C, wmOperator *op, const wmEvent
break;
}
default:
{
if ((event->val == KM_PRESS) && handleNumInput(C, &tgpi->num, event)) {
const float factor = tgpi->init_factor;
float value;
@ -649,6 +650,7 @@ static int gpencil_interpolate_modal(bContext *C, wmOperator *op, const wmEvent
/* unhandled event - allow to pass through */
return OPERATOR_RUNNING_MODAL | OPERATOR_PASS_THROUGH;
}
}
}
/* still running... */

38
source/blender/editors/interface/interface_handlers.c
View File

@ -3371,7 +3371,7 @@ static void ui_do_but_textedit(
if (event->type == WHEELDOWNMOUSE) {
break;
}
/* fall-through */
ATTR_FALLTHROUGH;
case ENDKEY:
ui_textedit_move(but, data, STRCUR_DIR_NEXT,
event->shift != 0, STRCUR_JUMP_ALL);
@ -3389,7 +3389,7 @@ static void ui_do_but_textedit(
if (event->type == WHEELUPMOUSE) {
break;
}
/* fall-through */
ATTR_FALLTHROUGH;
case HOMEKEY:
ui_textedit_move(but, data, STRCUR_DIR_PREV,
event->shift != 0, STRCUR_JUMP_ALL);
@ -8409,7 +8409,7 @@ static int ui_handle_button_event(bContext *C, const wmEvent *event, uiBut *but)
case MIDDLEMOUSE:
case MOUSEPAN:
UI_but_tooltip_timer_remove(C, but);
/* fall-through */
ATTR_FALLTHROUGH;
default:
/* handle button type specific events */
retval = ui_do_button(C, block, but, event);
@ -9203,23 +9203,23 @@ static int ui_handle_menu_event(
break;
case ONEKEY: case PAD1:
act = 1;
act = 1; ATTR_FALLTHROUGH;
case TWOKEY: case PAD2:
if (act == 0) act = 2;
if (act == 0) act = 2; ATTR_FALLTHROUGH;
case THREEKEY: case PAD3:
if (act == 0) act = 3;
if (act == 0) act = 3; ATTR_FALLTHROUGH;
case FOURKEY: case PAD4:
if (act == 0) act = 4;
if (act == 0) act = 4; ATTR_FALLTHROUGH;
case FIVEKEY: case PAD5:
if (act == 0) act = 5;
if (act == 0) act = 5; ATTR_FALLTHROUGH;
case SIXKEY: case PAD6:
if (act == 0) act = 6;
if (act == 0) act = 6; ATTR_FALLTHROUGH;
case SEVENKEY: case PAD7:
if (act == 0) act = 7;
if (act == 0) act = 7; ATTR_FALLTHROUGH;
case EIGHTKEY: case PAD8:
if (act == 0) act = 8;
if (act == 0) act = 8; ATTR_FALLTHROUGH;
case NINEKEY: case PAD9:
if (act == 0) act = 9;
if (act == 0) act = 9; ATTR_FALLTHROUGH;
case ZEROKEY: case PAD0:
if (act == 0) act = 10;
@ -9814,13 +9814,13 @@ static int ui_pie_handler(bContext *C, const wmEvent *event, uiPopupBlockHandle
case (ZEROKEY + n): case (PAD0 + n): \
{ if (num_dir == UI_RADIAL_NONE) num_dir = d; } (void)0
CASE_NUM_TO_DIR(1, UI_RADIAL_SW);
CASE_NUM_TO_DIR(2, UI_RADIAL_S);
CASE_NUM_TO_DIR(3, UI_RADIAL_SE);
CASE_NUM_TO_DIR(4, UI_RADIAL_W);
CASE_NUM_TO_DIR(6, UI_RADIAL_E);
CASE_NUM_TO_DIR(7, UI_RADIAL_NW);
CASE_NUM_TO_DIR(8, UI_RADIAL_N);
CASE_NUM_TO_DIR(1, UI_RADIAL_SW); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(2, UI_RADIAL_S); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(3, UI_RADIAL_SE); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(4, UI_RADIAL_W); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(6, UI_RADIAL_E); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(7, UI_RADIAL_NW); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(8, UI_RADIAL_N); ATTR_FALLTHROUGH;
CASE_NUM_TO_DIR(9, UI_RADIAL_NE);
{
but = ui_block_pie_dir_activate(block, event, num_dir);

4
source/blender/editors/mask/mask_ops.c
View File

@ -889,7 +889,7 @@ static int slide_point_modal(bContext *C, wmOperator *op, const wmEvent *event)
if (ELEM(event->type, LEFTSHIFTKEY, RIGHTSHIFTKEY))
data->is_accurate = (event->val == KM_PRESS);
/* fall-through */ /* update CV position */
ATTR_FALLTHROUGH; /* update CV position */
case MOUSEMOVE:
{
ScrArea *sa = CTX_wm_area(C);
@ -1376,7 +1376,7 @@ static int slide_spline_curvature_modal(bContext *C, wmOperator *op, const wmEve
}
/* fall-through */ /* update CV position */
ATTR_FALLTHROUGH; /* update CV position */
case MOUSEMOVE:
{
float B[2], mouse_coord[2], delta[2];

2
source/blender/editors/render/render_opengl.c
View File

@ -1070,7 +1070,7 @@ static int screen_opengl_render_modal(bContext *C, wmOperator *op, const wmEvent
/* render frame? */
if (oglrender->timer == event->customdata)
break;
/* fall-through */
ATTR_FALLTHROUGH;
default:
/* nothing to do */
return OPERATOR_RUNNING_MODAL;

3
source/blender/editors/screen/screen_edit.c
View File

@ -1863,10 +1863,11 @@ void ED_update_for_newframe(Main *bmain, Scene *scene, int UNUSED(mute))
/* update animated texture nodes */
{
Tex *tex;
for (tex = bmain->tex.first; tex; tex = tex->id.next)
for (tex = bmain->tex.first; tex; tex = tex->id.next) {
if (tex->use_nodes && tex->nodetree) {
ntreeTexTagAnimated(tex->nodetree);
}
}
}
}

2
source/blender/editors/space_clip/space_clip.c
View File

@ -337,7 +337,7 @@ static void clip_listener(bScreen *UNUSED(sc), ScrArea *sa, wmNotifier *wmn, con
switch (wmn->data) {
case ND_FRAME:
clip_scopes_tag_refresh(sa);
/* fall-through */
ATTR_FALLTHROUGH;
case ND_FRAME_RANGE:
ED_area_tag_redraw(sa);

3
source/blender/editors/space_clip/tracking_ops.c
View File

@ -875,8 +875,7 @@ static int slide_marker_modal(bContext *C, wmOperator *op, const wmEvent *event)
if (ELEM(event->type, LEFTSHIFTKEY, RIGHTSHIFTKEY)) {
data->accurate = event->val == KM_PRESS;
}
/* fall-through */
ATTR_FALLTHROUGH;
case MOUSEMOVE:
mdelta[0] = event->mval[0] - data->mval[0];
mdelta[1] = event->mval[1] - data->mval[1];

3
source/blender/editors/space_clip/tracking_ops_plane.c
View File

@ -301,8 +301,7 @@ static int slide_plane_marker_modal(bContext *C,
if (ELEM(event->type, LEFTSHIFTKEY, RIGHTSHIFTKEY)) {
data->accurate = event->val == KM_PRESS;
}
/* fall-through */
ATTR_FALLTHROUGH;
case MOUSEMOVE:
mdelta[0] = event->mval[0] - data->previous_mval[0];
mdelta[1] = event->mval[1] - data->previous_mval[1];

2
source/blender/editors/space_image/image_buttons.c
View File

@ -1208,7 +1208,7 @@ void uiTemplateImageStereo3d(uiLayout *layout, PointerRNA *stereo3d_format_ptr)
case S3D_DISPLAY_SIDEBYSIDE:
{
uiItemR(col, stereo3d_format_ptr, "use_sidebyside_crosseyed", 0, NULL, ICON_NONE);
/* fall-through */
ATTR_FALLTHROUGH;
}
case S3D_DISPLAY_TOPBOTTOM:
{

2
source/blender/editors/space_nla/nla_draw.c
View File

@ -391,7 +391,7 @@ static void nla_draw_strip(SpaceNla *snla, AnimData *adt, NlaTrack *nlt, NlaStri
/* draw the rect to the edge of the screen */
immRectf(shdr_pos, v2d->cur.xmin, yminc, strip->start, ymaxc);
}
/* fall-through */
ATTR_FALLTHROUGH;
/* this only draws after the strip */
case NLASTRIP_EXTEND_HOLD_FORWARD:

6
source/blender/editors/space_node/node_add.c
View File

@ -107,13 +107,15 @@ static bool add_reroute_intersect_check(bNodeLink *link, float mcoords[][2], int
if (node_link_bezier_points(NULL, NULL, link, coord_array, NODE_LINK_RESOL)) {
for (i = 0; i < tot - 1; i++)
for (b = 0; b < NODE_LINK_RESOL; b++)
for (i = 0; i < tot - 1; i++) {
for (b = 0; b < NODE_LINK_RESOL; b++) {
if (isect_seg_seg_v2(mcoords[i], mcoords[i + 1], coord_array[b], coord_array[b + 1]) > 0) {
result[0] = (mcoords[i][0] + mcoords[i + 1][0]) / 2.0f;
result[1] = (mcoords[i][1] + mcoords[i + 1][1]) / 2.0f;
return 1;
}
}
}
}
return 0;
}

2
source/blender/editors/space_sequencer/sequencer_edit.c
View File

@ -562,7 +562,7 @@ int seq_effect_find_selected(Scene *scene, Sequence *activeseq, int type, Sequen
}
if (seq1 == NULL) seq1 = seq2;
if (seq3 == NULL) seq3 = seq2;
/* fall-through */
ATTR_FALLTHROUGH;
case 2:
if (seq1 == NULL || seq2 == NULL) {
*error_str = N_("2 selected sequence strips are needed");

2
source/blender/editors/space_text/space_text.c
View File

@ -158,7 +158,7 @@ static void text_listener(bScreen *UNUSED(sc), ScrArea *sa, wmNotifier *wmn, con
}
ED_area_tag_redraw(sa);
/* fall-through */ /* fall down to tag redraw */
ATTR_FALLTHROUGH; /* fall down to tag redraw */
case NA_ADDED:
case NA_REMOVED:
ED_area_tag_redraw(sa);

4
source/blender/editors/space_text/text_autocomplete.c
View File

@ -455,7 +455,7 @@ static int text_autocomplete_modal(bContext *C, wmOperator *op, const wmEvent *e
break;
case PAGEDOWNKEY:
scroll = SUGG_LIST_SIZE - 1;
/* fall-through */
ATTR_FALLTHROUGH;
case WHEELDOWNMOUSE:
case DOWNARROWKEY:
if (event->val == KM_PRESS) {
@ -489,7 +489,7 @@ static int text_autocomplete_modal(bContext *C, wmOperator *op, const wmEvent *e
break;
case PAGEUPKEY:
scroll = SUGG_LIST_SIZE - 1;
/* fall-through */
ATTR_FALLTHROUGH;
case WHEELUPMOUSE:
case UPARROWKEY:
if (event->val == KM_PRESS) {

2
source/blender/editors/space_view3d/drawobject.c
View File

@ -6013,7 +6013,7 @@ static void draw_new_particle_system(Scene *scene, View3D *v3d, RegionView3D *rv
copy_m4_m4(imat, rv3d->viewinv);
normalize_v3(imat[0]);
normalize_v3(imat[1]);
/* fall-through */
ATTR_FALLTHROUGH;
case PART_DRAW_CROSS:
case PART_DRAW_AXIS:
/* lets calculate the scale: */

8
source/blender/editors/transform/transform.c
View File

@ -1616,7 +1616,7 @@ static void drawArrow(ArrowDirection d, short offset, short length, short size)
offset = -offset;
length = -length;
size = -size;
/* fall-through */
ATTR_FALLTHROUGH;
case RIGHT:
immVertex2f(POS_INDEX, offset, 0);
immVertex2f(POS_INDEX, offset + length, 0);
@ -1630,7 +1630,7 @@ static void drawArrow(ArrowDirection d, short offset, short length, short size)
offset = -offset;
length = -length;
size = -size;
/* fall-through */
ATTR_FALLTHROUGH;
case UP:
immVertex2f(POS_INDEX, 0, offset);
immVertex2f(POS_INDEX, 0, offset + length);
@ -1651,7 +1651,7 @@ static void drawArrowHead(ArrowDirection d, short size)
switch (d) {
case LEFT:
size = -size;
/* fall-through */
ATTR_FALLTHROUGH;
case RIGHT:
immVertex2f(POS_INDEX, 0, 0);
immVertex2f(POS_INDEX, -size, -size);
@ -1661,7 +1661,7 @@ static void drawArrowHead(ArrowDirection d, short size)
case DOWN:
size = -size;
/* fall-through */
ATTR_FALLTHROUGH;
case UP:
immVertex2f(POS_INDEX, 0, 0);
immVertex2f(POS_INDEX, -size, -size);

4
source/blender/editors/transform/transform_manipulator.c
View File

@ -949,7 +949,7 @@ static int calc_manipulator_stats(const bContext *C)
break;
}
/* if not gimbal, fall through to normal */
/* fall-through */
ATTR_FALLTHROUGH;
}
case V3D_MANIP_NORMAL:
{
@ -960,7 +960,7 @@ static int calc_manipulator_stats(const bContext *C)
break;
}
/* no break we define 'normal' as 'local' in Object mode */
/* fall-through */
ATTR_FALLTHROUGH;
}
case V3D_MANIP_LOCAL:
{

4
source/blender/editors/transform/transform_orientations.c
View File

@ -455,14 +455,14 @@ void initTransformOrientation(bContext *C, TransInfo *t)
BLI_strncpy(t->spacename, IFACE_("gimbal"), sizeof(t->spacename));
break;
}
/* fall-through */ /* no gimbal fallthrough to normal */
ATTR_FALLTHROUGH; /* no gimbal fallthrough to normal */
case V3D_MANIP_NORMAL:
if (obedit || (ob && ob->mode & OB_MODE_POSE)) {
BLI_strncpy(t->spacename, IFACE_("normal"), sizeof(t->spacename));
ED_getTransformOrientationMatrix(C, t->spacemtx, t->around);
break;
}
/* fall-through */ /* we define 'normal' as 'local' in Object mode */
ATTR_FALLTHROUGH; /* we define 'normal' as 'local' in Object mode */
case V3D_MANIP_LOCAL:
BLI_strncpy(t->spacename, IFACE_("local"), sizeof(t->spacename));

4
source/blender/editors/util/numinput.c
View File

@ -296,7 +296,7 @@ bool handleNumInput(bContext *C, NumInput *n, const wmEvent *event)
}
/* Else, common behavior with DELKEY, only difference is remove char(s) before/after the cursor. */
dir = STRCUR_DIR_PREV;
/* fall-through */
ATTR_FALLTHROUGH;
case DELKEY:
if ((n->val_flag[idx] & NUM_EDITED) && n->str[0]) {
int t_cur = cur = n->str_cur;
@ -322,7 +322,7 @@ bool handleNumInput(bContext *C, NumInput *n, const wmEvent *event)
break;
case LEFTARROWKEY:
dir = STRCUR_DIR_PREV;
/* fall-through */
ATTR_FALLTHROUGH;
case RIGHTARROWKEY:
cur = n->str_cur;
if (event->ctrl) {

1
source/blender/editors/uvedit/uvedit_smart_stitch.c
View File

@ -2173,6 +2173,7 @@ static int stitch_modal(bContext *C, wmOperator *op, const wmEvent *event)
}
break;
}
ATTR_FALLTHROUGH;
case PADENTER:
case RETKEY:
if (event->val == KM_PRESS) {

5
source/blender/gpu/intern/gpu_debug.c
View File

@ -29,6 +29,7 @@
* \ingroup gpu
*/
#include "BLI_compiler_attrs.h"
#include "BLI_utildefines.h"
#include "BLI_sys_types.h"
#include "BLI_system.h"
@ -94,7 +95,7 @@ static void APIENTRY gpu_debug_proc(
switch (severity) {
case GL_DEBUG_SEVERITY_HIGH:
backtrace = true;
/* fall through */
ATTR_FALLTHROUGH;
case GL_DEBUG_SEVERITY_MEDIUM:
case GL_DEBUG_SEVERITY_LOW:
case GL_DEBUG_SEVERITY_NOTIFICATION: /* KHR has this, ARB does not */
@ -134,7 +135,7 @@ static void APIENTRY gpu_debug_proc_amd(
switch (severity) {
case GL_DEBUG_SEVERITY_HIGH:
backtrace = true;
/* fall through */
ATTR_FALLTHROUGH;
case GL_DEBUG_SEVERITY_MEDIUM:
case GL_DEBUG_SEVERITY_LOW:
fprintf(stderr, "GL %s: %s\n", category_name_amd(category), message);

3
source/blender/imbuf/intern/png.c
View File

@ -613,11 +613,12 @@ ImBuf *imb_loadpng(const unsigned char *mem, size_t size, int flags, char colors
int unit_type;
png_uint_32 xres, yres;
if (png_get_pHYs(png_ptr, info_ptr, &xres, &yres, &unit_type))
if (png_get_pHYs(png_ptr, info_ptr, &xres, &yres, &unit_type)) {
if (unit_type == PNG_RESOLUTION_METER) {
ibuf->ppm[0] = xres;
ibuf->ppm[1] = yres;
}
}
}
}
else {

3
source/blender/imbuf/intern/radiance_hdr.c
View File

@ -340,13 +340,14 @@ static int fwritecolrs(FILE *file, int width, int channels, unsigned char *ibufs
}
if (((beg - j) > 1) && ((beg - j) < MINRUN)) {
c2 = j + 1;
while (rgbe_scan[c2++][i] == rgbe_scan[j][i])
while (rgbe_scan[c2++][i] == rgbe_scan[j][i]) {
if (c2 == beg) { /* short run */
putc((unsigned char)(128 + beg - j), file);
putc((unsigned char)(rgbe_scan[j][i]), file);
j = beg;
break;
}
}
}
while (j < beg) { /* write out non-run */
if ((c2 = beg - j) > 128) c2 = 128;

6
source/blender/python/intern/bpy_rna.c
View File

@ -806,7 +806,7 @@ static PyObject *pyrna_struct_richcmp(PyObject *a, PyObject *b, int op)
switch (op) {
case Py_NE:
ok = !ok;
/* fall-through */
ATTR_FALLTHROUGH;
case Py_EQ:
res = ok ? Py_False : Py_True;
break;
@ -836,7 +836,7 @@ static PyObject *pyrna_prop_richcmp(PyObject *a, PyObject *b, int op)
switch (op) {
case Py_NE:
ok = !ok;
/* fall-through */
ATTR_FALLTHROUGH;
case Py_EQ:
res = ok ? Py_False : Py_True;
break;
@ -5168,7 +5168,7 @@ static PyObject *pyrna_param_to_py(PointerRNA *ptr, PropertyRNA *prop, void *dat
ret = Matrix_CreatePyObject(data, 3, 3, NULL);
break;
}
/* fall-through */
ATTR_FALLTHROUGH;
#endif
default:
ret = PyTuple_New(len);

2
source/blender/python/mathutils/mathutils_Color.c
View File

@ -173,7 +173,7 @@ static PyObject *Color_richcmpr(PyObject *a, PyObject *b, int op)
switch (op) {
case Py_NE:
ok = !ok;
/* fall-through */
ATTR_FALLTHROUGH;
case Py_EQ:
res = ok ? Py_False : Py_True;
break;

4
source/blender/python/mathutils/mathutils_Euler.c
View File

@ -65,7 +65,7 @@ static PyObject *Euler_new(PyTypeObject *type, PyObject *args, PyObject *kwds)
case 2:
if ((order = euler_order_from_string(order_str, "mathutils.Euler()")) == -1)
return NULL;
/* fall-through */
ATTR_FALLTHROUGH;
case 1:
if (mathutils_array_parse(eul, EULER_SIZE, EULER_SIZE, seq, "mathutils.Euler()") == -1)
return NULL;
@ -370,7 +370,7 @@ static PyObject *Euler_richcmpr(PyObject *a, PyObject *b, int op)
switch (op) {
case Py_NE:
ok = !ok;
/* fall-through */
ATTR_FALLTHROUGH;
case Py_EQ:
res = ok ? Py_False : Py_True;
break;

2
source/blender/python/mathutils/mathutils_Matrix.c
View File

@ -2034,7 +2034,7 @@ static PyObject *Matrix_richcmpr(PyObject *a, PyObject *b, int op)
switch (op) {
case Py_NE:
ok = !ok;
/* fall-through */
ATTR_FALLTHROUGH;
case Py_EQ:
res = ok ? Py_False : Py_True;
break;

2
source/blender/python/mathutils/mathutils_Quaternion.c
View File

@ -552,7 +552,7 @@ static PyObject *Quaternion_richcmpr(PyObject *a, PyObject *b, int op)
switch (op) {
case Py_NE:
ok = !ok;
/* fall-through */
ATTR_FALLTHROUGH;
case Py_EQ:
res = ok ? Py_False : Py_True;
break;

9
source/blender/render/intern/source/convertblender.c
View File

@ -4702,6 +4702,12 @@ static void add_render_object(Render *re, Object *ob, Object *par, DupliObject *
if (ob->particlesystem.first) {
psysindex= 1;
for (psys=ob->particlesystem.first; psys; psys=psys->next, psysindex++) {
/* It seems that we may generate psys->renderdata recursively in some nasty intricated cases of
* several levels of bupliobject (see T51524).
* For now, basic rule is, do not restore psys if it was already in 'render state'.
* Another, more robust solution could be to add some reference counting to that renderdata... */
const bool psys_has_renderdata = (psys->renderdata != NULL);
if (!psys_check_enabled(ob, psys, G.is_rendering))
continue;
@ -4713,8 +4719,9 @@ static void add_render_object(Render *re, Object *ob, Object *par, DupliObject *
if (dob)
psys->flag |= PSYS_USE_IMAT;
init_render_object_data(re, obr, timeoffset);
if (!(re->r.scemode & R_VIEWPORT_PREVIEW))
if (!(re->r.scemode & R_VIEWPORT_PREVIEW) && !psys_has_renderdata) {
psys_render_restore(ob, psys);
}
psys->flag &= ~PSYS_USE_IMAT;
/* only add instance for objects that have not been used for dupli */

2
source/blender/render/intern/source/render_texture.c
View File

@ -1495,6 +1495,7 @@ void texture_rgb_blend(float in[3], const float tex[3], const float out[3], floa
case MTEX_SUB:
fact= -fact;
ATTR_FALLTHROUGH;
case MTEX_ADD:
fact*= facg;
in[0]= (fact*tex[0] + out[0]);
@ -1609,6 +1610,7 @@ float texture_value_blend(float tex, float out, float fact, float facg, int blen
case MTEX_SUB:
fact= -fact;
ATTR_FALLTHROUGH;
case MTEX_ADD:
in= fact*tex + out;
break;

6
source/blender/render/intern/source/rendercore.c
View File

@ -180,11 +180,12 @@ static void halo_pixelstruct(HaloRen *har, RenderLayer **rlpp, int totsample, in
flarec= 0;
if (fullsample) {
for (sample=0; sample<totsample; sample++)
for (sample=0; sample<totsample; sample++) {
if (ps->mask & (1 << sample)) {
float *pass = RE_RenderLayerGetPass(rlpp[sample], RE_PASSNAME_COMBINED, R.viewname);
addalphaAddfacFloat(pass + od*4, col, har->add);
}
}
}
else {
fac= ((float)amountm)/(float)R.osa;
@ -215,11 +216,12 @@ static void halo_pixelstruct(HaloRen *har, RenderLayer **rlpp, int totsample, in
}
if (fullsample) {
for (sample=0; sample<totsample; sample++)
for (sample=0; sample<totsample; sample++) {
if (!(mask & (1 << sample))) {
float *pass = RE_RenderLayerGetPass(rlpp[sample], RE_PASSNAME_COMBINED, R.viewname);
addalphaAddfacFloat(pass + od*4, col, har->add);
}
}
}
else {
col[0]= accol[0];

6
source/blender/render/intern/source/volumetric.c
View File

@ -275,10 +275,12 @@ static float metadensity(Object *ob, const float co[3])
break;
case MB_CUBE:
tp[2] = (tp[2] > ml->expz) ? (tp[2] - ml->expz) : ((tp[2] < -ml->expz) ? (tp[2] + ml->expz) : 0.f);
/* no break, xy as plane */
/* no break, xy as plane */
ATTR_FALLTHROUGH;
case MB_PLANE:
tp[1] = (tp[1] > ml->expy) ? (tp[1] - ml->expy) : ((tp[1] < -ml->expy) ? (tp[1] + ml->expy) : 0.f);
/* no break, x as tube */
/* no break, x as tube */
ATTR_FALLTHROUGH;
case MB_TUBE:
tp[0] = (tp[0] > ml->expx) ? (tp[0] - ml->expx) : ((tp[0] < -ml->expx) ? (tp[0] + ml->expx) : 0.f);
}

2
source/blender/render/intern/source/zbuf.c
View File

@ -3623,7 +3623,7 @@ static void merge_transp_passes(RenderLayer *rl, ShadeResult *shr)
fp[3] = speed[3];
}
}
}
}
}
if (col) {

8
source/gameengine/Expressions/intern/PyObjectPlus.cpp
View File

@ -335,7 +335,7 @@ PyObject *PyObjectPlus::py_get_attrdef(PyObject *self_py, const PyAttributeDef *
Py_DECREF(resultlist);
return NULL;
}
// walkthrough
ATTR_FALLTHROUGH;
case KX_PYATTRIBUTE_TYPE_INT:
{
int *val = reinterpret_cast<int*>(ptr);
@ -406,7 +406,7 @@ PyObject *PyObjectPlus::py_get_attrdef(PyObject *self_py, const PyAttributeDef *
{
return NULL;
}
// walkthrough
ATTR_FALLTHROUGH;
case KX_PYATTRIBUTE_TYPE_INT:
{
int *val = reinterpret_cast<int*>(ptr);
@ -638,7 +638,7 @@ int PyObjectPlus::py_set_attrdef(PyObject *self_py, PyObject *value, const PyAtt
PyErr_Format(PyExc_AttributeError, "Size check error for attribute, \"%s\", report to blender.org", attrdef->m_name);
goto UNDO_AND_ERROR;
}
// walkthrough
ATTR_FALLTHROUGH;
case KX_PYATTRIBUTE_TYPE_INT:
{
int *var = reinterpret_cast<int*>(ptr);
@ -883,7 +883,7 @@ int PyObjectPlus::py_set_attrdef(PyObject *self_py, PyObject *value, const PyAtt
PyErr_Format(PyExc_AttributeError, "attribute size check error for attribute \"%s\", report to blender.org", attrdef->m_name);
goto FREE_AND_ERROR;
}
// walkthrough
ATTR_FALLTHROUGH;
case KX_PYATTRIBUTE_TYPE_INT:
{
int *var = reinterpret_cast<int*>(ptr);

4
source/gameengine/GameLogic/SCA_PropertySensor.cpp
View File

@ -133,7 +133,7 @@ bool SCA_PropertySensor::CheckPropertyCondition()
{
case KX_PROPSENSOR_NOTEQUAL:
reverse = true;
/* fall-through */
ATTR_FALLTHROUGH;
case KX_PROPSENSOR_EQUAL:
{
CValue* orgprop = GetParent()->FindIdentifier(m_checkpropname);
@ -231,7 +231,7 @@ bool SCA_PropertySensor::CheckPropertyCondition()
}
case KX_PROPSENSOR_LESSTHAN:
reverse = true;
/* fall-through */
ATTR_FALLTHROUGH;
case KX_PROPSENSOR_GREATERTHAN:
{
CValue* orgprop = GetParent()->FindIdentifier(m_checkpropname);

1
source/gameengine/Ketsji/KX_ObstacleSimulation.cpp
View File

@ -345,6 +345,7 @@ static MT_Point3 nearestPointToObstacle(MT_Point3& pos ,KX_Obstacle* obstacle)
MT_Point3 res = obstacle->m_pos + abdir*proj;
return res;
}
ATTR_FALLTHROUGH;
}
case KX_OBSTACLE_CIRCLE :
default:

2
source/gameengine/Ketsji/KX_SoundActuator.cpp
View File

@ -108,7 +108,7 @@ void KX_SoundActuator::play()
case KX_SOUNDACT_LOOPBIDIRECTIONAL:
case KX_SOUNDACT_LOOPBIDIRECTIONAL_STOP:
sound = AUD_Sound_pingpong(sound);
// fall through
ATTR_FALLTHROUGH;
case KX_SOUNDACT_LOOPEND:
case KX_SOUNDACT_LOOPSTOP:
loop = true;