From 2ff9e5d46dcf23a010b8f6c219964ea726d1a145 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 --- hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp b/hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp index c817f4e42c..ed223efedf 100644 --- a/hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp +++ b/hipamd/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 da66236083cafe5203ad02f04aad9574d8400092 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 --- hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp | 2 +- 1 file changed, 1 insertion(+), 1 deletion(-) diff --git a/hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp index 322bd1370e..9842c17bc6 100644 --- a/hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/hipamd/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 2184e7f6da90613eb2ba8441ea519bd38391f9a7 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 --- hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp | 2 +- hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp | 1 - 2 files changed, 1 insertion(+), 2 deletions(-) diff --git a/hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp b/hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp index ed223efedf..8b237d7d24 100644 --- a/hipamd/samples/2_Cookbook/11_texture_driver/tex2dKernel.cpp +++ b/hipamd/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/hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp b/hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp index 9842c17bc6..2485d455ef 100644 --- a/hipamd/samples/2_Cookbook/11_texture_driver/texture2dDrv.cpp +++ b/hipamd/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;