UPGRADE YOUR BROWSER

We have detected your current browser version is not the latest one. Xilinx.com uses the latest web technologies to bring you the best online experience possible. Please upgrade to a Xilinx.com supported browser:Chrome, Firefox, Internet Explorer 11, Safari. Thank you!

cancel
Showing results for 
Search instead for 
Did you mean: 
Visitor motebphd2018
Visitor
203 Views
Registered: ‎05-31-2018

The xilinx on-chip global memory example failed in SDSoC 2018.1

Hi,

I am trying  to understand the use of on-chip glibal memory example that provided in  https://www.xilinx.com/support/documentation/sw_manuals/ug1207-sdaccel-performance-optimization.pdf (page 83).

However I got this error

ERROR: [KernelCheck 83-113] Kernel 'vadd' accesses non-pipe global variable 'xcl_mem_g_var1', which is not supported.

Error: unable to create function map for HLS kernel vadd

Does anyone have an idea on how to solve this error and get the example works?

--------------------------------------------------------------------

My host code is :

#include "xcl2.hpp"

#define LENGTH (1024)

int main(int argc, char* argv[])
{
size_t vector_size_bytes = sizeof(int) * LENGTH;
int t=1;
//Source Memories
std::vector<unsigned int> source_a(LENGTH);
std::vector<unsigned int> source_b(LENGTH);
std::vector<unsigned int> result_sim (LENGTH);
std::vector<unsigned int> result_krnl(LENGTH);

/* Create the test data and golden data locally */
for(int i=0; i < LENGTH; i++){
source_a[i] = 1;
source_b[i] = 1;
result_sim[i] = source_a[i] + source_b[i];
}

// OPENCL HOST CODE AREA START

//Getting Xilinx Platform and its device
std::vector<cl::Device> devices = xcl::get_xil_devices();
cl::Device device = devices[0];
std::string device_name = device.getInfo<CL_DEVICE_NAME>();

//Creating Context and Command Queue for selected Device
cl::Context context(device);
cl::CommandQueue q(context, device);

//Loading XCL Bin into char buffer
std::string binaryFile = xcl::find_binary_file(device_name,"vadd");
cl::Program::Binaries bins = xcl::import_binary_file(binaryFile);
devices.resize(1);
cl::Program program(context, devices, bins);

//Creating Kernel and Functor of Kernel
int err1;
cl::Kernel kernel_input(program, "vadd_dataIn", &err1);
auto krnl_in = cl::KernelFunctor<cl::Buffer&, cl::Buffer&>(kernel_input);
cl::Kernel kernel(program, "vadd", &err1);
auto krnl_vadd = cl::KernelFunctor<int>(kernel);
cl::Kernel kernel_output(program, "vadd_dataOut", &err1);
auto krnl_out = cl::KernelFunctor<cl::Buffer&>(kernel_output);

//Creating Buffers inside Device
cl::Buffer buffer_a(context, CL_MEM_READ_ONLY, vector_size_bytes);
cl::Buffer buffer_b(context, CL_MEM_READ_ONLY, vector_size_bytes);
cl::Buffer buffer_c(context, CL_MEM_WRITE_ONLY, vector_size_bytes);

//Copying input data to Device buffer from host memory
q.enqueueWriteBuffer(buffer_a, CL_TRUE, 0, vector_size_bytes, source_a.data());
q.enqueueWriteBuffer(buffer_b, CL_TRUE, 0, vector_size_bytes, source_b.data());

//Running Kernel
krnl_in (cl::EnqueueArgs(q, cl::NDRange(1,1,1), cl::NDRange(1,1,1)),
buffer_a, buffer_b);
krnl_vadd (cl::EnqueueArgs(q, cl::NDRange(1,1,1), cl::NDRange(1,1,1)), t);
krnl_out(cl::EnqueueArgs(q, cl::NDRange(1,1,1), cl::NDRange(1,1,1)),
buffer_c);
q.finish();

//Copying Device result data to Host memory
q.enqueueReadBuffer(buffer_c, CL_TRUE, 0, vector_size_bytes, result_krnl.data());

// OPENCL HOST CODE AREA END

/* Compare the results of the kernel to the simulation */
bool krnl_match = true;
for(int i = 0; i < LENGTH; i++){
if(result_sim[i] != result_krnl[i]){
std::cout <<"Error: Result mismatch" << std::endl;
std::cout <<"i = " << i << " CPU result = " << result_sim[i] << " Krnl Result = " << result_krnl[i] << std::endl;
krnl_match = false;
break;
}
}
std::cout << "TEST " << (krnl_match ? "PASSED" : "FAILED") << std::endl;
return (krnl_match ? EXIT_SUCCESS: EXIT_FAILURE);
}

--------------------------------------------------------------------

 

My kernel code is :

#define N 128

global int g_var0[1024];
global int g_var1[1024];
global int g_var2[1024];


__kernel void __attribute__((reqd_work_group_size(1,1,1)))
vadd_dataIn(
__global int* a,
__global int* b
){

for(int i=0; i<N; i++){
g_var0[i]= a[i];
g_var1[i]= b[i];
}
}

//------------------------------------------------------------------------------

__kernel void __attribute__ ((reqd_work_group_size(1, 1, 1)))
vadd( int inc) {

int input_data_a, input_data_b;
for(int i=0; i<N; i++){
input_data_a=g_var0[i];
input_data_b=g_var1[i];
g_var2[i]= input_data_a + input_data_b;
}
}

//------------------------------------------------------------------------------

__kernel void __attribute__((reqd_work_group_size(1,1,1)))
vadd_dataOut(
__global int* c
){
for(int i=0; i<N; i++){
c[i]= g_var2[i];
}
}

--------------------------------------------------------------------

 

The compilation output + the error :

make -j8 incremental
/home/motebphd2019/SDK/2018.1/gnu/aarch64/lin/aarch64-linux/bin/aarch64-linux-gnu-g++ -DSDX_PLATFORM=zcu102 -D__USE_XOPEN2K8 -I/home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/libs/xcl2 -I/home/motebphd2019/SDx/2018.1/runtime/include/1_2/ -I/home/motebphd2019/Vivado/2018.1/include/ -O2 -g -Wall -c -fmessage-length=0 -std=c++14 -DSDX_PLATFORM=zcu102 -D__USE_XOPEN2K8 -I/home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/libs/xcl2 -I/home/motebphd2019/SDx/2018.1/runtime/include/1_2/ -I/home/motebphd2019/Vivado/2018.1/include/ -O2 -g -Wall -c -fmessage-length=0 -o "src/host.o" "../src/host.cpp"
/home/motebphd2019/SDx/2018.1/bin/xocc -t hw --platform zcu102 --save-temps --report system --clkid 1 -c -k vadd --messageDb vadd/vadd.mdb -I"../src" --xp misc:solution_name=_xocc_compile_vadd_vadd -o"vadd/vadd.xo" "../src/vadd.cl"
/home/motebphd2019/SDx/2018.1/bin/xocc -t hw --platform zcu102 --save-temps --report system --clkid 1 -c -k vadd_dataOut --messageDb vadd/vadd_dataOut.mdb -I"../src" --xp misc:solution_name=_xocc_compile_vadd_vadd_dataOut -o"vadd/vadd_dataOut.xo" "../src/vadd.cl"
/home/motebphd2019/SDx/2018.1/bin/xocc -t hw --platform zcu102 --save-temps --report system --clkid 1 -c -k vadd_dataIn --messageDb vadd/vadd_dataIn.mdb -I"../src" --xp misc:solution_name=_xocc_compile_vadd_vadd_dataIn -o"vadd/vadd_dataIn.xo" "../src/vadd.cl"

****** xocc v2018.1 (64-bit)
**** SW Build 2188600 on Wed Apr 4 18:39:19 MDT 2018
** Copyright 1986-2018 Xilinx, Inc. All Rights Reserved.

Attempting to get a license: ap_opencl

****** xocc v2018.1 (64-bit)
**** SW Build 2188600 on Wed Apr 4 18:39:19 MDT 2018
** Copyright 1986-2018 Xilinx, Inc. All Rights Reserved.

Attempting to get a license: ap_opencl

****** xocc v2018.1 (64-bit)
**** SW Build 2188600 on Wed Apr 4 18:39:19 MDT 2018
** Copyright 1986-2018 Xilinx, Inc. All Rights Reserved.

Attempting to get a license: ap_opencl
WARNING: [XOCC 17-301] Failed to get a license for 'ap_opencl'. Explanation: The license feature ap_opencl could not be found.
Resolution: Check the status of your licenses in the Vivado License Manager. For debug help search Xilinx Support for "Licensing FAQ".
Attempting to get a license: ap_sdsoc
WARNING: [XOCC 17-301] Failed to get a license for 'ap_opencl'. Explanation: The license feature ap_opencl could not be found.
Resolution: Check the status of your licenses in the Vivado License Manager. For debug help search Xilinx Support for "Licensing FAQ".
Attempting to get a license: ap_sdsoc
INFO: [XOCC 17-1540] The version limit for your license is '2019.03' and has expired for new software. A version limit expiration means that, although you may be able to continue to use the current version of tools or IP with this license, you will not be eligible for any updates or new releases.
Feature available: ap_sdsoc
WARNING: [XOCC 17-301] Failed to get a license for 'ap_opencl'. Explanation: The license feature ap_opencl could not be found.
Resolution: Check the status of your licenses in the Vivado License Manager. For debug help search Xilinx Support for "Licensing FAQ".
Attempting to get a license: ap_sdsoc
INFO: [XOCC 17-1540] The version limit for your license is '2019.03' and has expired for new software. A version limit expiration means that, although you may be able to continue to use the current version of tools or IP with this license, you will not be eligible for any updates or new releases.
Feature available: ap_sdsoc
INFO: [XOCC 17-1540] The version limit for your license is '2019.03' and has expired for new software. A version limit expiration means that, although you may be able to continue to use the current version of tools or IP with this license, you will not be eligible for any updates or new releases.
Feature available: ap_sdsoc
INFO: [XOCC 60-585] Compiling for hardware target
INFO: [XOCC 60-585] Compiling for hardware target
INFO: [XOCC 60-585] Compiling for hardware target
INFO: [XOCC 60-895] Target platform: /home/motebphd2019/SDx/2018.1/platforms/zcu102/zcu102.xpfm
INFO: [XOCC 60-895] Target platform: /home/motebphd2019/SDx/2018.1/platforms/zcu102/zcu102.xpfm
INFO: [XOCC 60-895] Target platform: /home/motebphd2019/SDx/2018.1/platforms/zcu102/zcu102.xpfm
INFO: [XOCC 60-423] Target device: zcu102
INFO: [XOCC 60-242] Creating kernel: 'vadd_dataIn'
INFO: [XOCC 60-423] Target device: zcu102
INFO: [XOCC 60-242] Creating kernel: 'vadd_dataOut'
INFO: [XOCC 60-423] Target device: zcu102
INFO: [XOCC 60-242] Creating kernel: 'vadd'
/home/motebphd2019/SDK/2018.1/gnu/aarch64/lin/aarch64-linux/bin/aarch64-linux-gnu-g++ -o "L100-on-chip-global-mem.exe" libs/xcl2/xcl2.o src/host.o -lxilinxopencl -lpthread -lrt -ldl -lcrypt -lstdc++ -L/home/motebphd2019/SDx/2018.1/runtime/lib/aarch64 -lxilinxopencl -lpthread -lrt -ldl -lcrypt -lstdc++ -L/home/motebphd2019/SDx/2018.1/runtime/lib/aarch64
cp L100-on-chip-global-mem.exe sd_card/L100-on-chip-global-mem.exe

===>The following messages were generated while performing high-level synthesis for kernel: vadd_dataOut Log file:/home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/Debug/_xocc_compile_vadd_vadd_dataOut/impl/kernels/vadd_dataOut/vivado_hls.log :
INFO: [XOCC 204-61] Option 'relax_ii_for_timing' is enabled, will increase II to preserve clock frequency constraints.
INFO: [XOCC 204-61] Pipelining loop 'Loop 1'.
INFO: [XOCC 204-61] Pipelining result : Target II = 1, Final II = 1, Depth = 3.
INFO: [XOCC 60-594] Finished kernel compilation
INFO: [XOCC 60-244] Generating system estimate report...
INFO: [XOCC 60-677] Generated system_estimate.xtxt
INFO: [XOCC 60-586] Created vadd/vadd_dataOut.xo
INFO: [XOCC 60-791] Total elapsed time: 0h 0m 20s

===>The following messages were generated while performing high-level synthesis for kernel: vadd_dataIn Log file:/home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/Debug/_xocc_compile_vadd_vadd_dataIn/impl/kernels/vadd_dataIn/vivado_hls.log :
INFO: [XOCC 204-61] Option 'relax_ii_for_timing' is enabled, will increase II to preserve clock frequency constraints.
INFO: [XOCC 204-61] Pipelining loop 'Loop 1'.
INFO: [XOCC 204-61] Pipelining result : Target II = 1, Final II = 2, Depth = 138.
INFO: [XOCC 60-594] Finished kernel compilation
INFO: [XOCC 60-244] Generating system estimate report...
INFO: [XOCC 60-677] Generated system_estimate.xtxt
INFO: [XOCC 60-586] Created vadd/vadd_dataIn.xo
INFO: [XOCC 60-791] Total elapsed time: 0h 0m 21s

===>The following messages were generated while performing high-level synthesis for kernel: vadd Log file:/home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/Debug/_xocc_compile_vadd_vadd/impl/kernels/vadd/vivado_hls.log :
INFO: [XOCC 204-61] Option 'relax_ii_for_timing' is enabled, will increase II to preserve clock frequency constraints.
INFO: [XOCC 204-61] Pipelining loop 'Loop 1'.
INFO: [XOCC 204-61] Pipelining result : Target II = 1, Final II = 1, Depth = 4.
INFO: [XOCC 60-594] Finished kernel compilation
INFO: [XOCC 60-244] Generating system estimate report...
INFO: [XOCC 60-677] Generated system_estimate.xtxt
INFO: [XOCC 60-586] Created vadd/vadd.xo
INFO: [XOCC 60-791] Total elapsed time: 0h 0m 21s
/home/motebphd2019/SDx/2018.1/bin/xocc -t hw --platform zcu102 --save-temps --report system --clkid 1 -l --nk vadd:1 --nk vadd_dataIn:1 --nk vadd_dataOut:1 --messageDb vadd.mdb --xp misc:solution_name=_xocc_link_vadd --remote_ip_cache /home/motebphd2019/SD_workspace-2018.1/ip_cache -o"vadd.xclbin" vadd/vadd.xo vadd/vadd_dataIn.xo vadd/vadd_dataOut.xo --sys_config ocl

****** xocc v2018.1 (64-bit)
**** SW Build 2188600 on Wed Apr 4 18:39:19 MDT 2018
** Copyright 1986-2018 Xilinx, Inc. All Rights Reserved.

INFO: [XOCC 60-629] Linking for hardware target
INFO: [XOCC 60-895] Target platform: /home/motebphd2019/SDx/2018.1/platforms/zcu102/zcu102.xpfm
INFO: [XOCC 60-423] Target device: zcu102
INFO: [XOCC 60-825] xocc command line options for sdx_link are --xo vadd/vadd.xo --xo vadd/vadd_dataIn.xo --xo vadd/vadd_dataOut.xo --nk vadd:1 --nk vadd_dataIn:1 --nk vadd_dataOut:1 -clkid 1 -keep
INFO: [XOCC 60-824] sdx_link command line is sdx_link --xo vadd/vadd.xo --xo vadd/vadd_dataIn.xo --xo vadd/vadd_dataOut.xo --nk vadd:1 --nk vadd_dataIn:1 --nk vadd_dataOut:1 -clkid 1 -keep --xpfm /home/motebphd2019/SDx/2018.1/platforms/zcu102/zcu102.xpfm --target hw --output_dir . --temp_dir .
using /home/motebphd2019/SDx/2018.1/platforms/zcu102/zcu102.xpfm
extracting xo v3 file /home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/Debug/vadd/vadd.xo
ERROR: [KernelCheck 83-113] Kernel 'vadd' accesses non-pipe global variable 'xcl_mem_g_var1', which is not supported.

Error generating intermediate file /home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/Debug/iprepo/temp/xo0/ip_repo/xilinx_com_hls_vadd_1_0/vadd.fcnmap.xml
Error: unable to create function map for HLS kernel vadd
Unable to process /home/motebphd2019/SD_workspace-2018.1/L100-on-chip-global-mem/Debug/vadd/vadd.xo
Error processing .xo files, exiting
ERROR: [XOCC 60-398] sdx_link failed
ERROR: [XOCC 60-626] Kernel link failed to complete
ERROR: [XOCC 60-703] Failed to finish linking
makefile:92: recipe for target 'vadd.xclbin' failed
make: *** [vadd.xclbin] Error 1

 

0 Kudos