From a1b608bf0d323b82d43964da71ba43de8da2efd3 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Fri, 16 Mar 2018 12:35:25 +0530 Subject: [PATCH 1/3] Fixed function not found issue --- samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp b/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp index c817f4e42c..ed223efedf 100644 --- a/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp +++ b/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp @@ -23,7 +23,7 @@ THE SOFTWARE. #include "hip/hip_runtime.h" extern texture tex; -__global__ void tex2dKernel(hipLaunchParm lp, float* outputData, int width, int height) { +extern "C" __global__ void tex2dKernel(hipLaunchParm lp, float* outputData, int width, int height) { int x = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x; int y = hipBlockIdx_y * hipBlockDim_y + hipThreadIdx_y; outputData[y * width + x] = tex2D(tex, x, y); From be25556aedf3f7cbf7a4a0b6dc2a7ea637ce1174 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Fri, 16 Mar 2018 12:54:44 +0530 Subject: [PATCH 2/3] Change co file name --- samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp index 322bd1370e..9842c17bc6 100644 --- a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp @@ -27,7 +27,7 @@ THE SOFTWARE. #include #include -#define fileName "tex2dKernel.code" +#define fileName "tex2dKernel.code.adipose" texture tex; bool testResult = false; From ed2d4ddfc7cb0dc1efcc0ba75c9e82e065157863 Mon Sep 17 00:00:00 2001 From: Rahul Garg Date: Fri, 16 Mar 2018 22:50:25 +0530 Subject: [PATCH 3/3] Removed hidden args and hipLaunchParm from HIP/HCC path --- samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp | 2 +- samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp | 1 - 2 files changed, 1 insertion(+), 2 deletions(-) diff --git a/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp b/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp index ed223efedf..8b237d7d24 100644 --- a/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp +++ b/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp @@ -23,7 +23,7 @@ THE SOFTWARE. #include "hip/hip_runtime.h" extern texture tex; -extern "C" __global__ void tex2dKernel(hipLaunchParm lp, float* outputData, int width, int height) { +extern "C" __global__ void tex2dKernel(float* outputData, int width, int height) { int x = hipBlockIdx_x * hipBlockDim_x + hipThreadIdx_x; int y = hipBlockIdx_y * hipBlockDim_y + hipThreadIdx_y; outputData[y * width + x] = tex2D(tex, x, y); diff --git a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp index 9842c17bc6..2485d455ef 100644 --- a/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp @@ -90,7 +90,6 @@ bool runTest(int argc, char** argv) { #ifdef __HIP_PLATFORM_HCC__ struct { - uint32_t _hidden[6]; // genco path + wrapper-gen pass used hidden arguments. void* _Ad; unsigned int _Bd; unsigned int _Cd;