Intel® Quartus® Prime Software
Intel® Quartus® Prime Design Software, Design Entry, Synthesis, Simulation, Verification, Timing Analysis, System Design (Platform Designer, formerly Qsys)
Announcements
FPGA community forums and blogs on community.intel.com are migrating to the new Altera Community and are read-only. For urgent support needs during this transition, please visit the FPGA Design Resources page or contact an Altera Authorized Distributor.
17267 Discussions

Compiler Error: Multiple channel write sites.

Altera_Forum
Honored Contributor II
1,779 Views

Hi, 

 

My code as follows: 

 

#ifndef SIMD_WORK_ITEMS 

#define SIMD_WORK_ITEMS 4 // default value 

#endif 

 

 

#pragma OPENCL_EXTENSION cl_altera_channels : enable 

 

 

// Channel declarations 

channel int DATA __attribute__((depth(8))); 

 

 

__kernel 

__attribute((reqd_work_group_size(BLOCK_SIZE,BLOCK_SIZE,1))) 

__attribute((num_simd_work_items(SIMD_WORK_ITEMS))) 

void matrixMult( __global int *restrict A, __global int *restrict B) 

// Write result to channel 

int off = get_global_id(1) * get_global_size(0) + get_global_id(0); 

int running_sum = A[off] * B[off]; 

write_channel_altera(DATA, running_sum); 

 

 

__kernel 

void FindMinAndMax(int size, __global int *restrict data) 

int min_temp = 0; 

int max_temp = 0; 

 

 

for (int i = 0; i < size; i++) { 

int val = read_channel_altera(DATA); 

min_temp = min(min_temp, val); 

max_temp = max(max_temp, val); 

 

 

data[i] = val; 

 

 

When I compile the kernels with command "aoc device/matrix_mult_bak.cl -o bin/matrix_mult_bak.aocx --report -v", it print out errors  

"Compiler Error: Multiple channel write sites.". What the error means?How to solve the problem?Thank you!
0 Kudos
2 Replies
Altera_Forum
Honored Contributor II
947 Views

I have solved the problem, just delete "__attribute((num_simd_work_items(SIMD_WORK_ITEMS)) )" .

0 Kudos
Altera_Forum
Honored Contributor II
947 Views

For future reference, "Compiler Error: Multiple channel write sites." means that you are writing to the same channel from multiple locations in your code, which is not allowed. Each channel can only have one read and one write port. When you use SIMD, the compiler will try to extend the kernel pipeline to run multiple work-items in parallel, and if you have a channel read or write in your code, this will result in multiple work-items trying to read from or write to the same channel at the same time which is not allowed. SIMD and kernel pipeline replication (num_compute_units) are not supported for NDRange kernels that have channels.

0 Kudos
Reply