Этот коммит содержится в:
Aditya Atluri
2016-04-22 09:14:26 -05:00
родитель 091ec702a2
Коммит a2bc6ebc9f
2 изменённых файлов: 309 добавлений и 0 удалений
+14
Просмотреть файл
@@ -21,10 +21,16 @@ THE SOFTWARE.
#define HIP_LDG_H
#if __HCC__
<<<<<<< HEAD
#include"hip_vector_types.h"
#include"host_defines.h"
#if __hcc_workweek__ >= 16164
=======
#include"hip/hip_vector_types.h"
#include"hip/hcc_detail/host_defines.h"
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
__device__ char __ldg(const char* );
__device__ char1 __ldg(const char1* );
__device__ char2 __ldg(const char2* );
@@ -94,6 +100,14 @@ __device__ double2 __ldg(const double2* );
__device__ double3 __ldg(const double3* );
__device__ double4 __ldg(const double4* );
<<<<<<< HEAD
#endif // __hcc_workweek__
#endif // __HCC__
#endif // HIP_LDG_H
=======
#endif
#endif
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
+295
Просмотреть файл
@@ -18,6 +18,128 @@ THE SOFTWARE.
*/
#include"hcc_detail/hip_ldg.h"
<<<<<<< HEAD
#if __hcc_workweek__ >= 16164
__device__ char __ldg(const char* ptr)
{
return ptr[0];
}
__device__ char1 __ldg(const char1* ptr)
{
return ptr[0];
}
__device__ char2 __ldg(const char2* ptr)
{
return ptr[0];
}
__device__ char3 __ldg(const char3* ptr)
{
return ptr[0];
}
__device__ char4 __ldg(const char4* ptr)
{
return ptr[0];
}
__device__ signed char __ldg(const signed char* ptr)
{
return ptr[0];
}
__device__ unsigned char __ldg(const unsigned char* ptr)
{
return ptr[0];
}
__device__ short __ldg(const short* ptr)
{
return ptr[0];
}
__device__ short1 __ldg(const short1* ptr)
{
return ptr[0];
}
__device__ short2 __ldg(const short2* ptr)
{
return ptr[0];
}
__device__ short3 __ldg(const short3* ptr)
{
return ptr[0];
}
__device__ short4 __ldg(const short4* ptr)
{
return ptr[0];
}
__device__ unsigned short __ldg(const unsigned short* ptr)
{
return ptr[0];
}
__device__ int __ldg(const int* ptr)
{
return ptr[0];
}
__device__ int1 __ldg(const int1* ptr)
{
return ptr[0];
}
__device__ int2 __ldg(const int2* ptr)
{
return ptr[0];
}
__device__ int3 __ldg(const int3* ptr)
{
return ptr[0];
}
__device__ int4 __ldg(const int4* ptr)
{
return ptr[0];
}
__device__ unsigned int __ldg(const unsigned int* ptr)
{
return ptr[0];
}
__device__ long __ldg(const long* ptr)
{
return ptr[0];
}
__device__ long1 __ldg(const long1* ptr)
{
return ptr[0];
}
__device__ long2 __ldg(const long2* ptr)
{
return ptr[0];
}
__device__ long3 __ldg(const long3* ptr)
{
return ptr[0];
}
__device__ long4 __ldg(const long4* ptr)
=======
__device__ char __ldg(const char* ptr)
{
@@ -25,37 +147,210 @@ __device__ char __ldg(const char* ptr)
}
__device__ signed char __ldg(const signed char* ptr)
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
{
return ptr[0];
}
<<<<<<< HEAD
__device__ unsigned long __ldg(const unsigned long* ptr)
=======
__device__ short __ldg(const short* ptr)
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
{
return ptr[0];
}
<<<<<<< HEAD
__device__ long long __ldg(const long long* ptr)
=======
__device__ int __ldg(const int* ptr)
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
{
return ptr[0];
}
<<<<<<< HEAD
__device__ longlong1 __ldg(const longlong1* ptr)
=======
__device__ long long __ldg(const long long* ptr)
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
{
return ptr[0];
}
<<<<<<< HEAD
__device__ longlong2 __ldg(const longlong2* ptr)
{
return ptr[0];
}
__device__ longlong3 __ldg(const longlong3* ptr)
=======
__device__ int2 __ldg(const int2* ptr)
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
{
return ptr[0];
}
<<<<<<< HEAD
__device__ longlong4 __ldg(const longlong4* ptr)
{
return ptr[0];
}
__device__ unsigned long long __ldg(const unsigned long long* ptr)
{
return ptr[0];
}
__device__ uchar1 __ldg(const uchar1* ptr)
=======
__device__ int4 __ldg(const int4* ptr)
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4
{
return ptr[0];
}
<<<<<<< HEAD
__device__ uchar2 __ldg(const uchar2* ptr)
{
return ptr[0];
}
__device__ uchar3 __ldg(const uchar3* ptr)
{
return ptr[0];
}
__device__ uchar4 __ldg(const uchar4* ptr)
{
return ptr[0];
}
__device__ ushort1 __ldg(const ushort1* ptr)
{
return ptr[0];
}
__device__ ushort2 __ldg(const ushort2* ptr)
{
return ptr[0];
}
__device__ ushort3 __ldg(const ushort3* ptr)
{
return ptr[0];
}
__device__ ushort4 __ldg(const ushort4* ptr)
{
return ptr[0];
}
__device__ uint1 __ldg(const uint1* ptr)
{
return ptr[0];
}
__device__ uint2 __ldg(const uint2* ptr)
{
return ptr[0];
}
__device__ uint3 __ldg(const uint3* ptr)
{
return ptr[0];
}
__device__ uint4 __ldg(const uint4* ptr)
{
return ptr[0];
}
__device__ ulonglong1 __ldg(const ulonglong1* ptr)
{
return ptr[0];
}
__device__ ulonglong2 __ldg(const ulonglong2* ptr)
{
return ptr[0];
}
__device__ ulonglong3 __ldg(const ulonglong3* ptr)
{
return ptr[0];
}
__device__ ulonglong4 __ldg(const ulonglong4* ptr)
{
return ptr[0];
}
__device__ float __ldg(const float* ptr)
{
return ptr[0];
}
__device__ float1 __ldg(const float1* ptr)
{
return ptr[0];
}
__device__ float2 __ldg(const float2* ptr)
{
return ptr[0];
}
__device__ float3 __ldg(const float3* ptr)
{
return ptr[0];
}
__device__ float4 __ldg(const float4* ptr)
{
return ptr[0];
}
__device__ double __ldg(const double* ptr)
{
return ptr[0];
}
__device__ double1 __ldg(const double1* ptr)
{
return ptr[0];
}
__device__ double2 __ldg(const double2* ptr)
{
return ptr[0];
}
__device__ double3 __ldg(const double3* ptr)
{
return ptr[0];
}
__device__ double4 __ldg(const double4* ptr)
{
return ptr[0];
}
#endif
=======
__device__ float __ldg(const float* ptr)
{
return ptr[0];
}
>>>>>>> 75532471b2d124c4af823e838bf983a20debe6c4