Merge branch 'master' into blender2.8

Conflicts:
	source/blender/editors/gpencil/drawgpencil.c

4 files changed, 34 insertions, 21 deletions

diff --git a/intern/cycles/device/opencl/opencl_util.cpp b/intern/cycles/device/opencl/opencl_util.cpp
index 6dca642f3f3..fe1c65a2224 100644
--- a/intern/cycles/device/opencl/opencl_util.cpp
+++ b/intern/cycles/device/opencl/opencl_util.cpp
@@ -1058,13 +1058,16 @@ cl_device_type OpenCLInfo::get_device_type(cl_device_id device_id)
 string OpenCLInfo::get_readable_device_name(cl_device_id device_id)
 {
 	char board_name[1024];
+	size_t length = 0;
 	if(clGetDeviceInfo(device_id,
 	                   CL_DEVICE_BOARD_NAME_AMD,
 	                   sizeof(board_name),
 	                   &board_name,
-	                   NULL) == CL_SUCCESS)
+					   &length) == CL_SUCCESS)
 	{
-		return board_name;
+		if(length != 0 && board_name[0] != '\0') {
+			return board_name;
+		}
 	}
 	/* Fallback to standard device name API. */
 	return get_device_name(device_id);
diff --git a/intern/cycles/kernel/geom/geom_curve.h b/intern/cycles/kernel/geom/geom_curve.h
index bb33b91847e..8888000f0e6 100644
--- a/intern/cycles/kernel/geom/geom_curve.h
+++ b/intern/cycles/kernel/geom/geom_curve.h
@@ -270,7 +270,7 @@ ccl_device_curveintersect bool bvh_cardinal_curve_intersect(KernelGlobals *kg, I
 		int ka = max(k0 - 1, v00.x);
 		int kb = min(k1 + 1, v00.x + v00.y - 1);
 
-#if defined(__KERNEL_AVX2__) && (!defined(_MSC_VER) || _MSC_VER > 1800)
+#if defined(__KERNEL_AVX2__) && defined(__KERNEL_SSE__) && (!defined(_MSC_VER) || _MSC_VER > 1800)
 		avxf P_curve_0_1, P_curve_2_3;
 		if(is_curve_primitive) {
 			P_curve_0_1 = _mm256_loadu2_m128(&kg->__curve_keys.data[k0].x, &kg->__curve_keys.data[ka].x);
@@ -305,7 +305,7 @@ ccl_device_curveintersect bool bvh_cardinal_curve_intersect(KernelGlobals *kg, I
 		ssef htfm1 = shuffle<1, 0, 1, 3>(load1f_first(extract<0>(d_ss)), vdir0);
 		ssef htfm2 = shuffle<1, 3, 2, 3>(mul_shuf, vdir0);
 
-#if defined(__KERNEL_AVX2__) && (!defined(_MSC_VER) || _MSC_VER > 1800)
+#if defined(__KERNEL_AVX2__) && defined(__KERNEL_SSE__) && (!defined(_MSC_VER) || _MSC_VER > 1800)
 		const avxf vPP = _mm256_broadcast_ps(&P.m128);
 		const avxf htfm00 = avxf(htfm0.m128, htfm0.m128);
 		const avxf htfm11 = avxf(htfm1.m128, htfm1.m128);
diff --git a/intern/cycles/kernel/geom/geom_motion_curve.h b/intern/cycles/kernel/geom/geom_motion_curve.h
index dc1388b6643..119bdb2f15c 100644
--- a/intern/cycles/kernel/geom/geom_motion_curve.h
+++ b/intern/cycles/kernel/geom/geom_motion_curve.h
@@ -152,7 +152,7 @@ ccl_device_inline void motion_cardinal_curve_keys(KernelGlobals *kg,
 	keys[3] = (1.0f - t)*keys[3] + t*next_keys[3];
 }
 
-#ifdef __KERNEL_AVX2__
+#if defined(__KERNEL_AVX2__) && defined(__KERNEL_SSE__)
 /* Similar to above, but returns keys as pair of two AVX registers with each
  * holding two float4.
  */
diff --git a/intern/cycles/kernel/kernel_compat_cpu.h b/intern/cycles/kernel/kernel_compat_cpu.h
index cad5f4d2959..21da180bb8e 100644
--- a/intern/cycles/kernel/kernel_compat_cpu.h
+++ b/intern/cycles/kernel/kernel_compat_cpu.h
@@ -353,7 +353,7 @@ template<typename T> struct texture_image  {
 	{
 		int ix, iy, iz;
 		int nix, niy, niz;
-		
+
 		float tx = frac(x*(float)width - 0.5f, &ix);
 		float ty = frac(y*(float)height - 0.5f, &iy);
 		float tz = frac(z*(float)depth - 0.5f, &iz);
@@ -404,7 +404,18 @@ template<typename T> struct texture_image  {
 		return r;
 	}
 
-	ccl_never_inline float4 interp_3d_ex_tricubic(float x, float y, float z)
+	/* TODO(sergey): For some unspeakable reason both GCC-6 and Clang-3.9 are
+	 * causing stack overflow issue in this function unless it is inlined.
+	 *
+	 * Only happens for AVX2 kernel and global __KERNEL_SSE__ vectorization
+	 * enabled.
+	 */
+#ifdef __GNUC__
+	ccl_always_inline
+#else
+	ccl_never_inline
+#endif
+	float4 interp_3d_ex_tricubic(float x, float y, float z)
 	{
 		int ix, iy, iz;
 		int nix, niy, niz;
@@ -463,13 +474,13 @@ template<typename T> struct texture_image  {
 
 		const int xc[4] = {pix, ix, nix, nnix};
 		const int yc[4] = {width * piy,
-						   width * iy,
-						   width * niy,
-						   width * nniy};
+		                   width * iy,
+		                   width * niy,
+		                   width * nniy};
 		const int zc[4] = {width * height * piz,
-						   width * height * iz,
-						   width * height * niz,
-						   width * height * nniz};
+		                   width * height * iz,
+		                   width * height * niz,
+		                   width * height * nniz};
 		float u[4], v[4], w[4];
 
 		/* Some helper macro to keep code reasonable size,
@@ -478,14 +489,14 @@ template<typename T> struct texture_image  {
 #define DATA(x, y, z) (read(data[xc[x] + yc[y] + zc[z]]))
 #define COL_TERM(col, row) \
 		(v[col] * (u[0] * DATA(0, col, row) + \
-				   u[1] * DATA(1, col, row) + \
-				   u[2] * DATA(2, col, row) + \
-				   u[3] * DATA(3, col, row)))
+		           u[1] * DATA(1, col, row) + \
+		           u[2] * DATA(2, col, row) + \
+		           u[3] * DATA(3, col, row)))
 #define ROW_TERM(row) \
 		(w[row] * (COL_TERM(0, row) + \
-				   COL_TERM(1, row) + \
-				   COL_TERM(2, row) + \
-				   COL_TERM(3, row)))
+		           COL_TERM(1, row) + \
+		           COL_TERM(2, row) + \
+		           COL_TERM(3, row)))
 
 		SET_CUBIC_SPLINE_WEIGHTS(u, tx);
 		SET_CUBIC_SPLINE_WEIGHTS(v, ty);
@@ -502,11 +513,10 @@ template<typename T> struct texture_image  {
 	ccl_always_inline float4 interp_3d_ex(float x, float y, float z,
 	                                      int interpolation = INTERPOLATION_LINEAR)
 	{
-		
 		if(UNLIKELY(!data))
 			return make_float4(0.0f, 0.0f, 0.0f, 0.0f);
 
-		switch(interpolation) { 
+		switch(interpolation) {
 			case INTERPOLATION_CLOSEST:
 				return interp_3d_ex_closest(x, y, z);
 			case INTERPOLATION_LINEAR: