blender
/
blender
129
398
Fork 572
Code Issues 6.2k Pull Requests 746 Projects 19 Wiki Activity

Cycles: Remove fermi related defines from the code. 

Did not touch Texture related defines, that comes next.
This commit is contained in:
Thomas Dinges 2018-02-17 22:19:54 +01:00
parent 2eaf90b305
commit e1ef902058
4 changed files with 5 additions and 31 deletions
Show Stats Download Patch File Download Diff File Expand all files Collapse all files

10
intern/cycles/kernel/geom/geom_curve_intersect.h
View File

@ -18,12 +18,6 @@ CCL_NAMESPACE_BEGIN
#ifdef __HAIR__
#if defined(__KERNEL_CUDA__) && (__CUDA_ARCH__ < 300)
# define ccl_device_curveintersect ccl_device
#else
# define ccl_device_curveintersect ccl_device_forceinline
#endif
#ifdef __KERNEL_SSE2__
ccl_device_inline ssef transform_point_T3(const ssef t[3], const ssef &a)
{
@ -32,7 +26,7 @@ ccl_device_inline ssef transform_point_T3(const ssef t[3], const ssef &a)
#endif
/* On CPU pass P and dir by reference to aligned vector. */
ccl_device_curveintersect bool cardinal_curve_intersect(
ccl_device_forceinline bool cardinal_curve_intersect(
KernelGlobals *kg,
Intersection *isect,
const float3 ccl_ref P,
@ -505,7 +499,7 @@ ccl_device_curveintersect bool cardinal_curve_intersect(
return hit;
}
ccl_device_curveintersect bool curve_intersect(KernelGlobals *kg,
ccl_device_forceinline bool curve_intersect(KernelGlobals *kg,
Intersection *isect,
float3 P,
float3 direction,

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

@ -50,10 +50,7 @@ __device__ half __float2half(const float f)
/* Qualifier wrappers for different names on different devices */
#define ccl_device __device__ __inline__
#if __CUDA_ARCH__ < 300
# define ccl_device_inline __device__ __inline__
# define ccl_device_forceinline __device__ __forceinline__
#elif __CUDA_ARCH__ < 500
#if __CUDA_ARCH__ < 500
# define ccl_device_inline __device__ __forceinline__
# define ccl_device_forceinline __device__ __forceinline__
#else

14
intern/cycles/kernel/kernels/cuda/kernel_config.h
View File

@ -16,20 +16,8 @@
/* device data taken from CUDA occupancy calculator */
/* 2.0 and 2.1 */
#if __CUDA_ARCH__ == 200 || __CUDA_ARCH__ == 210
# define CUDA_MULTIPRESSOR_MAX_REGISTERS 32768
# define CUDA_MULTIPROCESSOR_MAX_BLOCKS 8
# define CUDA_BLOCK_MAX_THREADS 1024
# define CUDA_THREAD_MAX_REGISTERS 63
/* tunable parameters */
# define CUDA_THREADS_BLOCK_WIDTH 16
# define CUDA_KERNEL_MAX_REGISTERS 32
# define CUDA_KERNEL_BRANCHED_MAX_REGISTERS 40
/* 3.0 and 3.5 */
#elif __CUDA_ARCH__ == 300 || __CUDA_ARCH__ == 350
#if __CUDA_ARCH__ == 300 || __CUDA_ARCH__ == 350
# define CUDA_MULTIPRESSOR_MAX_REGISTERS 65536
# define CUDA_MULTIPROCESSOR_MAX_BLOCKS 16
# define CUDA_BLOCK_MAX_THREADS 1024

7
intern/cycles/util/util_math_intersect.h
View File

@ -79,12 +79,7 @@ ccl_device bool ray_aligned_disk_intersect(
return true;
}
#if defined(__KERNEL_CUDA__) && __CUDA_ARCH__ < 300
ccl_device_inline
#else
ccl_device_forceinline
#endif
bool ray_triangle_intersect(
ccl_device_forceinline bool ray_triangle_intersect(
float3 ray_P, float3 ray_dir, float ray_t,
#if defined(__KERNEL_SSE2__) && defined(__KERNEL_SSE__)
const ssef *ssef_verts,